diff options
Diffstat (limited to 'llvm/test')
44 files changed, 1469 insertions, 1231 deletions
diff --git a/llvm/test/Bitcode/thinlto-alias-addrspacecast.ll b/llvm/test/Bitcode/thinlto-alias-addrspacecast.ll new file mode 100644 index 0000000..fe4f05e --- /dev/null +++ b/llvm/test/Bitcode/thinlto-alias-addrspacecast.ll @@ -0,0 +1,7 @@ +; RUN: opt -module-summary < %s | llvm-dis | FileCheck %s + +@__oclc_ABI_version = linkonce_odr hidden addrspace(4) constant i32 500, align 4 +@_ZL20__oclc_ABI_version__ = internal alias i32, addrspacecast (ptr addrspace(4) @__oclc_ABI_version to ptr) + +; CHECK: ^1 = gv: (name: "__oclc_ABI_version", summaries: (variable: (module: ^0, flags: {{.*}}))) +; CHECK: ^2 = gv: (name: "_ZL20__oclc_ABI_version__", summaries: (alias: (module: ^0, flags: {{.*}}, aliasee: ^1))) diff --git a/llvm/test/CodeGen/AArch64/spill-fill-zpr-predicates.mir b/llvm/test/CodeGen/AArch64/spill-fill-zpr-predicates.mir deleted file mode 100644 index 0298168..0000000 --- a/llvm/test/CodeGen/AArch64/spill-fill-zpr-predicates.mir +++ /dev/null @@ -1,1009 +0,0 @@ -# NOTE: Assertions have been autogenerated by utils/update_mir_test_checks.py UTC_ARGS: --version 5 -# RUN: llc -mtriple=aarch64-linux-gnu -aarch64-enable-zpr-predicate-spills -run-pass=greedy %s -o - | FileCheck %s -# RUN: llc -mtriple=aarch64-linux-gnu -aarch64-enable-zpr-predicate-spills -start-before=greedy -stop-after=aarch64-expand-pseudo -verify-machineinstrs %s -o - | FileCheck %s --check-prefix=EXPAND ---- | - source_filename = "<stdin>" - target datalayout = "e-m:e-i8:8:32-i16:16:32-i64:64-i128:128-n32:64-S128" - target triple = "aarch64--linux-gnu" - - define aarch64_sve_vector_pcs void @zpr_predicate_spill() #0 { entry: unreachable } - - define aarch64_sve_vector_pcs void @zpr_predicate_spill__save_restore_nzcv() #0 { entry: unreachable } - - define aarch64_sve_vector_pcs void @zpr_predicate_spill__save_restore_nzcv__scavenge_csr_gpr() #0 { entry: unreachable } - - define aarch64_sve_vector_pcs void @zpr_predicate_spill__spill_zpr() #0 { entry: unreachable } - - define aarch64_sve_vector_pcs void @zpr_predicate_spill_above_p7() #0 { entry: unreachable } - - define aarch64_sve_vector_pcs void @zpr_predicate_spill_p4_saved() #0 { entry: unreachable } - - attributes #0 = {nounwind "target-features"="+sme,+sve" "aarch64_pstate_sm_compatible"} -... ---- -name: zpr_predicate_spill -tracksRegLiveness: true -stack: -liveins: - - { reg: '$p0' } -body: | - bb.0.entry: - liveins: $p0 - - ; CHECK-LABEL: name: zpr_predicate_spill - ; CHECK: stack: - ; CHECK: - { id: 0, name: '', type: spill-slot, offset: 0, size: 16, alignment: 16, - ; CHECK-NEXT: stack-id: scalable-vector, callee-saved-register: - ; CHECK: liveins: $p0 - ; CHECK-NEXT: {{ $}} - ; - ; CHECK-NEXT: SPILL_PPR_TO_ZPR_SLOT_PSEUDO $p0, %stack.0, 0 :: (store (s128) into %stack.0) - ; - ; CHECK-NEXT: $p0 = IMPLICIT_DEF - ; CHECK-NEXT: $p1 = IMPLICIT_DEF - ; CHECK-NEXT: $p2 = IMPLICIT_DEF - ; CHECK-NEXT: $p3 = IMPLICIT_DEF - ; CHECK-NEXT: $p4 = IMPLICIT_DEF - ; CHECK-NEXT: $p5 = IMPLICIT_DEF - ; CHECK-NEXT: $p6 = IMPLICIT_DEF - ; CHECK-NEXT: $p7 = IMPLICIT_DEF - ; CHECK-NEXT: $p8 = IMPLICIT_DEF - ; CHECK-NEXT: $p9 = IMPLICIT_DEF - ; CHECK-NEXT: $p10 = IMPLICIT_DEF - ; CHECK-NEXT: $p11 = IMPLICIT_DEF - ; CHECK-NEXT: $p12 = IMPLICIT_DEF - ; CHECK-NEXT: $p13 = IMPLICIT_DEF - ; CHECK-NEXT: $p14 = IMPLICIT_DEF - ; CHECK-NEXT: $p15 = IMPLICIT_DEF - ; - ; CHECK-NEXT: $p0 = FILL_PPR_FROM_ZPR_SLOT_PSEUDO %stack.0, 0 :: (load (s128) from %stack.0) - ; - ; CHECK-NEXT: RET_ReallyLR implicit $p0 - - ; EXPAND-LABEL: name: zpr_predicate_spill - ; EXPAND: liveins: $p0, $fp, $p15, $p14, $p13, $p12, $p11, $p10, $p9, $p8, $p7, $p6, $p5, $p4 - ; EXPAND-NEXT: {{ $}} - ; - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1040, 0 - ; EXPAND-NEXT: frame-setup STRXui killed $fp, $sp, 128 :: (store (s64) into %stack.14) - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -12, implicit $vg - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p15, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 0 :: (store (s128) into %stack.13) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p14, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 1 :: (store (s128) into %stack.12) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p13, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 2 :: (store (s128) into %stack.11) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p12, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 3 :: (store (s128) into %stack.10) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p11, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 4 :: (store (s128) into %stack.9) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p10, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 5 :: (store (s128) into %stack.8) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p9, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 6 :: (store (s128) into %stack.7) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p8, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 7 :: (store (s128) into %stack.6) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p7, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 8 :: (store (s128) into %stack.5) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p6, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 9 :: (store (s128) into %stack.4) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p5, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 10 :: (store (s128) into %stack.3) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p4, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 11 :: (store (s128) into %stack.2) - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -1, implicit $vg - ; - ; EXPAND-NEXT: $z0 = CPY_ZPzI_B $p0, 1, 0 - ; EXPAND-NEXT: $x8 = ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: STR_ZXI $z0, $x8, 0 :: (store (s128) into %stack.0) - ; - ; EXPAND-NEXT: $p0 = IMPLICIT_DEF - ; EXPAND-NEXT: $p1 = IMPLICIT_DEF - ; EXPAND-NEXT: $p2 = IMPLICIT_DEF - ; EXPAND-NEXT: $p3 = IMPLICIT_DEF - ; EXPAND-NEXT: $p4 = IMPLICIT_DEF - ; EXPAND-NEXT: $p5 = IMPLICIT_DEF - ; EXPAND-NEXT: $p6 = IMPLICIT_DEF - ; EXPAND-NEXT: $p7 = IMPLICIT_DEF - ; EXPAND-NEXT: $p8 = IMPLICIT_DEF - ; EXPAND-NEXT: $p9 = IMPLICIT_DEF - ; EXPAND-NEXT: $p10 = IMPLICIT_DEF - ; EXPAND-NEXT: $p11 = IMPLICIT_DEF - ; EXPAND-NEXT: $p12 = IMPLICIT_DEF - ; EXPAND-NEXT: $p13 = IMPLICIT_DEF - ; EXPAND-NEXT: $p14 = IMPLICIT_DEF - ; EXPAND-NEXT: $p15 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = LDR_ZXI killed $x8, 0 :: (load (s128) from %stack.0) - ; EXPAND-NEXT: $p1 = frame-destroy PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p0 = CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 1, implicit $vg - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 0 :: (load (s128) from %stack.13) - ; EXPAND-NEXT: $p15 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 1 :: (load (s128) from %stack.12) - ; EXPAND-NEXT: $p14 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 2 :: (load (s128) from %stack.11) - ; EXPAND-NEXT: $p13 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 3 :: (load (s128) from %stack.10) - ; EXPAND-NEXT: $p12 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 4 :: (load (s128) from %stack.9) - ; EXPAND-NEXT: $p11 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 5 :: (load (s128) from %stack.8) - ; EXPAND-NEXT: $p10 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 6 :: (load (s128) from %stack.7) - ; EXPAND-NEXT: $p9 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 7 :: (load (s128) from %stack.6) - ; EXPAND-NEXT: $p8 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 8 :: (load (s128) from %stack.5) - ; EXPAND-NEXT: $p7 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 9 :: (load (s128) from %stack.4) - ; EXPAND-NEXT: $p6 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 10 :: (load (s128) from %stack.3) - ; EXPAND-NEXT: $p5 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 11 :: (load (s128) from %stack.2) - ; EXPAND-NEXT: $p4 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 12, implicit $vg - ; EXPAND-NEXT: $fp = frame-destroy LDRXui $sp, 128 :: (load (s64) from %stack.14) - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1040, 0 - ; EXPAND-NEXT: RET undef $lr, implicit $p0 - %1:ppr = COPY $p0 - - $p0 = IMPLICIT_DEF - $p1 = IMPLICIT_DEF - $p2 = IMPLICIT_DEF - $p3 = IMPLICIT_DEF - $p4 = IMPLICIT_DEF - $p5 = IMPLICIT_DEF - $p6 = IMPLICIT_DEF - $p7 = IMPLICIT_DEF - $p8 = IMPLICIT_DEF - $p9 = IMPLICIT_DEF - $p10 = IMPLICIT_DEF - $p11 = IMPLICIT_DEF - $p12 = IMPLICIT_DEF - $p13 = IMPLICIT_DEF - $p14 = IMPLICIT_DEF - $p15 = IMPLICIT_DEF - - $p0 = COPY %1 - - RET_ReallyLR implicit $p0 -... ---- -name: zpr_predicate_spill__save_restore_nzcv -tracksRegLiveness: true -stack: -liveins: - - { reg: '$p0' } -body: | - bb.0.entry: - liveins: $p0 - - ; CHECK-LABEL: name: zpr_predicate_spill__save_restore_nzcv - ; CHECK: stack: - ; CHECK: - { id: 0, name: '', type: spill-slot, offset: 0, size: 16, alignment: 16, - ; CHECK-NEXT: stack-id: scalable-vector, callee-saved-register: - ; CHECK: liveins: $p0 - ; CHECK-NEXT: {{ $}} - ; - ; CHECK-NEXT: $nzcv = IMPLICIT_DEF - ; - ; CHECK-NEXT: SPILL_PPR_TO_ZPR_SLOT_PSEUDO $p0, %stack.0, 0 :: (store (s128) into %stack.0) - ; - ; CHECK-NEXT: $p0 = IMPLICIT_DEF - ; CHECK-NEXT: $p1 = IMPLICIT_DEF - ; CHECK-NEXT: $p2 = IMPLICIT_DEF - ; CHECK-NEXT: $p3 = IMPLICIT_DEF - ; CHECK-NEXT: $p4 = IMPLICIT_DEF - ; CHECK-NEXT: $p5 = IMPLICIT_DEF - ; CHECK-NEXT: $p6 = IMPLICIT_DEF - ; CHECK-NEXT: $p7 = IMPLICIT_DEF - ; CHECK-NEXT: $p8 = IMPLICIT_DEF - ; CHECK-NEXT: $p9 = IMPLICIT_DEF - ; CHECK-NEXT: $p10 = IMPLICIT_DEF - ; CHECK-NEXT: $p11 = IMPLICIT_DEF - ; CHECK-NEXT: $p12 = IMPLICIT_DEF - ; CHECK-NEXT: $p13 = IMPLICIT_DEF - ; CHECK-NEXT: $p14 = IMPLICIT_DEF - ; CHECK-NEXT: $p15 = IMPLICIT_DEF - ; - ; CHECK-NEXT: $p0 = FILL_PPR_FROM_ZPR_SLOT_PSEUDO %stack.0, 0 :: (load (s128) from %stack.0) - ; - ; CHECK-NEXT: FAKE_USE implicit $nzcv - ; - ; CHECK-NEXT: RET_ReallyLR implicit $p0 - - ; EXPAND-LABEL: name: zpr_predicate_spill__save_restore_nzcv - ; EXPAND: liveins: $p0, $fp, $p15, $p14, $p13, $p12, $p11, $p10, $p9, $p8, $p7, $p6, $p5, $p4 - ; EXPAND-NEXT: {{ $}} - ; - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1040, 0 - ; EXPAND-NEXT: frame-setup STRXui killed $fp, $sp, 128 :: (store (s64) into %stack.14) - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -12, implicit $vg - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p15, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 0 :: (store (s128) into %stack.13) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p14, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 1 :: (store (s128) into %stack.12) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p13, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 2 :: (store (s128) into %stack.11) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p12, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 3 :: (store (s128) into %stack.10) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p11, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 4 :: (store (s128) into %stack.9) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p10, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 5 :: (store (s128) into %stack.8) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p9, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 6 :: (store (s128) into %stack.7) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p8, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 7 :: (store (s128) into %stack.6) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p7, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 8 :: (store (s128) into %stack.5) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p6, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 9 :: (store (s128) into %stack.4) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p5, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 10 :: (store (s128) into %stack.3) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p4, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 11 :: (store (s128) into %stack.2) - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -1, implicit $vg - ; - ; EXPAND-NEXT: $nzcv = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = CPY_ZPzI_B $p0, 1, 0 - ; EXPAND-NEXT: $x8 = ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: STR_ZXI $z0, $x8, 0 :: (store (s128) into %stack.0) - ; - ; EXPAND-NEXT: $p0 = IMPLICIT_DEF - ; EXPAND-NEXT: $p1 = IMPLICIT_DEF - ; EXPAND-NEXT: $p2 = IMPLICIT_DEF - ; EXPAND-NEXT: $p3 = IMPLICIT_DEF - ; EXPAND-NEXT: $p4 = IMPLICIT_DEF - ; EXPAND-NEXT: $p5 = IMPLICIT_DEF - ; EXPAND-NEXT: $p6 = IMPLICIT_DEF - ; EXPAND-NEXT: $p7 = IMPLICIT_DEF - ; EXPAND-NEXT: $p8 = IMPLICIT_DEF - ; EXPAND-NEXT: $p9 = IMPLICIT_DEF - ; EXPAND-NEXT: $p10 = IMPLICIT_DEF - ; EXPAND-NEXT: $p11 = IMPLICIT_DEF - ; EXPAND-NEXT: $p12 = IMPLICIT_DEF - ; EXPAND-NEXT: $p13 = IMPLICIT_DEF - ; EXPAND-NEXT: $p14 = IMPLICIT_DEF - ; EXPAND-NEXT: $p15 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = LDR_ZXI killed $x8, 0 :: (load (s128) from %stack.0) - ; EXPAND-NEXT: $fp = MRS 55824, implicit-def $nzcv, implicit $nzcv - ; EXPAND-NEXT: $p0 = PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p0 = CMPNE_PPzZI_B $p0, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: MSR 55824, $fp, implicit-def $nzcv - ; - ; EXPAND-NEXT: FAKE_USE implicit $nzcv - ; - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 1, implicit $vg - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 0 :: (load (s128) from %stack.13) - ; EXPAND-NEXT: $p1 = frame-destroy PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p15 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 1 :: (load (s128) from %stack.12) - ; EXPAND-NEXT: $p14 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 2 :: (load (s128) from %stack.11) - ; EXPAND-NEXT: $p13 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 3 :: (load (s128) from %stack.10) - ; EXPAND-NEXT: $p12 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 4 :: (load (s128) from %stack.9) - ; EXPAND-NEXT: $p11 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 5 :: (load (s128) from %stack.8) - ; EXPAND-NEXT: $p10 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 6 :: (load (s128) from %stack.7) - ; EXPAND-NEXT: $p9 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 7 :: (load (s128) from %stack.6) - ; EXPAND-NEXT: $p8 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 8 :: (load (s128) from %stack.5) - ; EXPAND-NEXT: $p7 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 9 :: (load (s128) from %stack.4) - ; EXPAND-NEXT: $p6 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 10 :: (load (s128) from %stack.3) - ; EXPAND-NEXT: $p5 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 11 :: (load (s128) from %stack.2) - ; EXPAND-NEXT: $p4 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 12, implicit $vg - ; EXPAND-NEXT: $fp = frame-destroy LDRXui $sp, 128 :: (load (s64) from %stack.14) - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1040, 0 - ; EXPAND-NEXT: RET undef $lr, implicit $p0 - $nzcv = IMPLICIT_DEF - - %1:ppr = COPY $p0 - - $p0 = IMPLICIT_DEF - $p1 = IMPLICIT_DEF - $p2 = IMPLICIT_DEF - $p3 = IMPLICIT_DEF - $p4 = IMPLICIT_DEF - $p5 = IMPLICIT_DEF - $p6 = IMPLICIT_DEF - $p7 = IMPLICIT_DEF - $p8 = IMPLICIT_DEF - $p9 = IMPLICIT_DEF - $p10 = IMPLICIT_DEF - $p11 = IMPLICIT_DEF - $p12 = IMPLICIT_DEF - $p13 = IMPLICIT_DEF - $p14 = IMPLICIT_DEF - $p15 = IMPLICIT_DEF - - $p0 = COPY %1 - - FAKE_USE implicit $nzcv - - RET_ReallyLR implicit $p0 -... ---- -name: zpr_predicate_spill__save_restore_nzcv__scavenge_csr_gpr -tracksRegLiveness: true -stack: -liveins: - - { reg: '$p0' } - - { reg: '$x0' } - - { reg: '$x1' } - - { reg: '$x2' } - - { reg: '$x3' } - - { reg: '$x4' } - - { reg: '$x5' } - - { reg: '$x6' } - - { reg: '$x7' } -body: | - bb.0.entry: - liveins: $p0, $x0, $x1, $x2, $x3, $x4, $x5, $x6, $x7 - - ; CHECK-LABEL: name: zpr_predicate_spill__save_restore_nzcv__scavenge_csr_gpr - ; CHECK: stack: - ; CHECK: - { id: 0, name: '', type: spill-slot, offset: 0, size: 16, alignment: 16, - ; CHECK-NEXT: stack-id: scalable-vector, callee-saved-register: - ; CHECK: liveins: $p0, $x0, $x1, $x2, $x3, $x4, $x5, $x6, $x7 - ; CHECK-NEXT: {{ $}} - ; - ; CHECK-NEXT: $nzcv = IMPLICIT_DEF - ; - ; CHECK-NEXT: $x8 = IMPLICIT_DEF - ; CHECK-NEXT: $x9 = IMPLICIT_DEF - ; CHECK-NEXT: $x10 = IMPLICIT_DEF - ; CHECK-NEXT: $x11 = IMPLICIT_DEF - ; CHECK-NEXT: $x12 = IMPLICIT_DEF - ; CHECK-NEXT: $x13 = IMPLICIT_DEF - ; CHECK-NEXT: $x14 = IMPLICIT_DEF - ; CHECK-NEXT: $x15 = IMPLICIT_DEF - ; CHECK-NEXT: $x16 = IMPLICIT_DEF - ; CHECK-NEXT: $x17 = IMPLICIT_DEF - ; CHECK-NEXT: $x18 = IMPLICIT_DEF - ; - ; CHECK-NEXT: SPILL_PPR_TO_ZPR_SLOT_PSEUDO $p0, %stack.0, 0 :: (store (s128) into %stack.0) - ; - ; CHECK-NEXT: $p0 = IMPLICIT_DEF - ; CHECK-NEXT: $p1 = IMPLICIT_DEF - ; CHECK-NEXT: $p2 = IMPLICIT_DEF - ; CHECK-NEXT: $p3 = IMPLICIT_DEF - ; CHECK-NEXT: $p4 = IMPLICIT_DEF - ; CHECK-NEXT: $p5 = IMPLICIT_DEF - ; CHECK-NEXT: $p6 = IMPLICIT_DEF - ; CHECK-NEXT: $p7 = IMPLICIT_DEF - ; CHECK-NEXT: $p8 = IMPLICIT_DEF - ; CHECK-NEXT: $p9 = IMPLICIT_DEF - ; CHECK-NEXT: $p10 = IMPLICIT_DEF - ; CHECK-NEXT: $p11 = IMPLICIT_DEF - ; CHECK-NEXT: $p12 = IMPLICIT_DEF - ; CHECK-NEXT: $p13 = IMPLICIT_DEF - ; CHECK-NEXT: $p14 = IMPLICIT_DEF - ; CHECK-NEXT: $p15 = IMPLICIT_DEF - ; - ; CHECK-NEXT: $p0 = FILL_PPR_FROM_ZPR_SLOT_PSEUDO %stack.0, 0 :: (load (s128) from %stack.0) - ; - ; CHECK-NEXT: FAKE_USE implicit $nzcv, implicit $x8, implicit $x9, implicit $x10, implicit $x11, implicit $x12, implicit $x13, implicit $x14, implicit $x15, implicit $x16, implicit $x17, implicit $x18 - ; - ; CHECK-NEXT: RET_ReallyLR implicit $p0, implicit $x0, implicit $x1, implicit $x2, implicit $x3, implicit $x4, implicit $x5, implicit $x6, implicit $x7, implicit $x8, implicit $x9, implicit $x10, implicit $x11, implicit $x12, implicit $x13, implicit $x14, implicit $x15, implicit $x16, implicit $x17, implicit $x18 - - ; EXPAND-LABEL: name: zpr_predicate_spill__save_restore_nzcv__scavenge_csr_gpr - ; EXPAND: liveins: $p0, $x0, $x1, $x2, $x3, $x4, $x5, $x6, $x7, $fp, $p15, $p14, $p13, $p12, $p11, $p10, $p9, $p8, $p7, $p6, $p5, $p4 - ; EXPAND-NEXT: {{ $}} - ; - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1040, 0 - ; EXPAND-NEXT: frame-setup STRXui killed $fp, $sp, 128 :: (store (s64) into %stack.14) - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -12, implicit $vg - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p15, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 0 :: (store (s128) into %stack.13) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p14, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 1 :: (store (s128) into %stack.12) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p13, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 2 :: (store (s128) into %stack.11) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p12, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 3 :: (store (s128) into %stack.10) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p11, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 4 :: (store (s128) into %stack.9) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p10, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 5 :: (store (s128) into %stack.8) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p9, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 6 :: (store (s128) into %stack.7) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p8, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 7 :: (store (s128) into %stack.6) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p7, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 8 :: (store (s128) into %stack.5) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p6, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 9 :: (store (s128) into %stack.4) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p5, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 10 :: (store (s128) into %stack.3) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p4, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 11 :: (store (s128) into %stack.2) - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -1, implicit $vg - ; - ; EXPAND-NEXT: $nzcv = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $x8 = IMPLICIT_DEF - ; EXPAND-NEXT: $x9 = IMPLICIT_DEF - ; EXPAND-NEXT: $x10 = IMPLICIT_DEF - ; EXPAND-NEXT: $x11 = IMPLICIT_DEF - ; EXPAND-NEXT: $x12 = IMPLICIT_DEF - ; EXPAND-NEXT: $x13 = IMPLICIT_DEF - ; EXPAND-NEXT: $x14 = IMPLICIT_DEF - ; EXPAND-NEXT: $x15 = IMPLICIT_DEF - ; EXPAND-NEXT: $x16 = IMPLICIT_DEF - ; EXPAND-NEXT: $x17 = IMPLICIT_DEF - ; EXPAND-NEXT: $x18 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = CPY_ZPzI_B $p0, 1, 0 - ; EXPAND-NEXT: $fp = ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: STR_ZXI $z0, $fp, 0 :: (store (s128) into %stack.0) - ; - ; EXPAND-NEXT: $p0 = IMPLICIT_DEF - ; EXPAND-NEXT: $p1 = IMPLICIT_DEF - ; EXPAND-NEXT: $p2 = IMPLICIT_DEF - ; EXPAND-NEXT: $p3 = IMPLICIT_DEF - ; EXPAND-NEXT: $p4 = IMPLICIT_DEF - ; EXPAND-NEXT: $p5 = IMPLICIT_DEF - ; EXPAND-NEXT: $p6 = IMPLICIT_DEF - ; EXPAND-NEXT: $p7 = IMPLICIT_DEF - ; EXPAND-NEXT: $p8 = IMPLICIT_DEF - ; EXPAND-NEXT: $p9 = IMPLICIT_DEF - ; EXPAND-NEXT: $p10 = IMPLICIT_DEF - ; EXPAND-NEXT: $p11 = IMPLICIT_DEF - ; EXPAND-NEXT: $p12 = IMPLICIT_DEF - ; EXPAND-NEXT: $p13 = IMPLICIT_DEF - ; EXPAND-NEXT: $p14 = IMPLICIT_DEF - ; EXPAND-NEXT: $p15 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = LDR_ZXI killed $fp, 0 :: (load (s128) from %stack.0) - ; EXPAND-NEXT: $fp = MRS 55824, implicit-def $nzcv, implicit $nzcv - ; EXPAND-NEXT: $p0 = PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p0 = CMPNE_PPzZI_B $p0, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: MSR 55824, $fp, implicit-def $nzcv - ; - ; EXPAND-NEXT: FAKE_USE implicit $nzcv, implicit $x8, implicit $x9, implicit $x10, implicit $x11, implicit $x12, implicit $x13, implicit $x14, implicit $x15, implicit $x16, implicit $x17, implicit $x18 - ; - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 1, implicit $vg - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 0 :: (load (s128) from %stack.13) - ; EXPAND-NEXT: $p1 = frame-destroy PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p15 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 1 :: (load (s128) from %stack.12) - ; EXPAND-NEXT: $p14 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 2 :: (load (s128) from %stack.11) - ; EXPAND-NEXT: $p13 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 3 :: (load (s128) from %stack.10) - ; EXPAND-NEXT: $p12 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 4 :: (load (s128) from %stack.9) - ; EXPAND-NEXT: $p11 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 5 :: (load (s128) from %stack.8) - ; EXPAND-NEXT: $p10 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 6 :: (load (s128) from %stack.7) - ; EXPAND-NEXT: $p9 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 7 :: (load (s128) from %stack.6) - ; EXPAND-NEXT: $p8 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 8 :: (load (s128) from %stack.5) - ; EXPAND-NEXT: $p7 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 9 :: (load (s128) from %stack.4) - ; EXPAND-NEXT: $p6 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 10 :: (load (s128) from %stack.3) - ; EXPAND-NEXT: $p5 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 11 :: (load (s128) from %stack.2) - ; EXPAND-NEXT: $p4 = frame-destroy CMPNE_PPzZI_B $p1, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 12, implicit $vg - ; EXPAND-NEXT: $fp = frame-destroy LDRXui $sp, 128 :: (load (s64) from %stack.14) - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1040, 0 - ; EXPAND-NEXT: RET undef $lr, implicit $p0, implicit $x0, implicit $x1, implicit $x2, implicit $x3, implicit $x4, implicit $x5, implicit $x6, implicit $x7, implicit $x8, implicit $x9, implicit $x10, implicit $x11, implicit $x12, implicit $x13, implicit $x14, implicit $x15, implicit $x16, implicit $x17, implicit $x18 - $nzcv = IMPLICIT_DEF - $x8 = IMPLICIT_DEF - $x9 = IMPLICIT_DEF - $x10 = IMPLICIT_DEF - $x11 = IMPLICIT_DEF - $x12 = IMPLICIT_DEF - $x13 = IMPLICIT_DEF - $x14 = IMPLICIT_DEF - $x15 = IMPLICIT_DEF - $x16 = IMPLICIT_DEF - $x17 = IMPLICIT_DEF - $x18 = IMPLICIT_DEF - - %1:ppr = COPY $p0 - - $p0 = IMPLICIT_DEF - $p1 = IMPLICIT_DEF - $p2 = IMPLICIT_DEF - $p3 = IMPLICIT_DEF - $p4 = IMPLICIT_DEF - $p5 = IMPLICIT_DEF - $p6 = IMPLICIT_DEF - $p7 = IMPLICIT_DEF - $p8 = IMPLICIT_DEF - $p9 = IMPLICIT_DEF - $p10 = IMPLICIT_DEF - $p11 = IMPLICIT_DEF - $p12 = IMPLICIT_DEF - $p13 = IMPLICIT_DEF - $p14 = IMPLICIT_DEF - $p15 = IMPLICIT_DEF - - $p0 = COPY %1 - - FAKE_USE implicit $nzcv, implicit $x8, implicit $x9, implicit $x10, implicit $x11, implicit $x12, implicit $x13, implicit $x14, implicit $x15, implicit $x16, implicit $x17, implicit $x18 - - RET_ReallyLR implicit $p0, implicit $x0, implicit $x1, implicit $x2, implicit $x3, implicit $x4, implicit $x5, implicit $x6, implicit $x7, implicit $x8, implicit $x9, implicit $x10, implicit $x11, implicit $x12, implicit $x13, implicit $x14, implicit $x15, implicit $x16, implicit $x17, implicit $x18 -... ---- -name: zpr_predicate_spill__spill_zpr -tracksRegLiveness: true -stack: -liveins: - - { reg: '$p0' } - - { reg: '$z0' } - - { reg: '$z1' } - - { reg: '$z2' } - - { reg: '$z3' } - - { reg: '$z4' } - - { reg: '$z5' } - - { reg: '$z6' } - - { reg: '$z7' } -body: | - bb.0.entry: - liveins: $p0, $z0, $z1, $z2, $z3, $z4, $z5, $z6, $z7 - - ; CHECK-LABEL: name: zpr_predicate_spill__spill_zpr - ; CHECK: stack: - ; CHECK: - { id: 0, name: '', type: spill-slot, offset: 0, size: 16, alignment: 16, - ; CHECK-NEXT: stack-id: scalable-vector, callee-saved-register: - ; CHECK: liveins: $p0, $z0, $z1, $z2, $z3, $z4, $z5, $z6, $z7 - ; CHECK-NEXT: {{ $}} - ; - ; CHECK-NEXT: $z16 = IMPLICIT_DEF - ; CHECK-NEXT: $z17 = IMPLICIT_DEF - ; CHECK-NEXT: $z18 = IMPLICIT_DEF - ; CHECK-NEXT: $z19 = IMPLICIT_DEF - ; CHECK-NEXT: $z20 = IMPLICIT_DEF - ; CHECK-NEXT: $z21 = IMPLICIT_DEF - ; CHECK-NEXT: $z22 = IMPLICIT_DEF - ; CHECK-NEXT: $z23 = IMPLICIT_DEF - ; CHECK-NEXT: $z24 = IMPLICIT_DEF - ; CHECK-NEXT: $z25 = IMPLICIT_DEF - ; CHECK-NEXT: $z26 = IMPLICIT_DEF - ; CHECK-NEXT: $z27 = IMPLICIT_DEF - ; CHECK-NEXT: $z28 = IMPLICIT_DEF - ; CHECK-NEXT: $z29 = IMPLICIT_DEF - ; CHECK-NEXT: $z30 = IMPLICIT_DEF - ; CHECK-NEXT: $z31 = IMPLICIT_DEF - ; - ; CHECK-NEXT: SPILL_PPR_TO_ZPR_SLOT_PSEUDO $p0, %stack.0, 0 :: (store (s128) into %stack.0) - ; - ; CHECK-NEXT: $p0 = IMPLICIT_DEF - ; CHECK-NEXT: $p1 = IMPLICIT_DEF - ; CHECK-NEXT: $p2 = IMPLICIT_DEF - ; CHECK-NEXT: $p3 = IMPLICIT_DEF - ; CHECK-NEXT: $p4 = IMPLICIT_DEF - ; CHECK-NEXT: $p5 = IMPLICIT_DEF - ; CHECK-NEXT: $p6 = IMPLICIT_DEF - ; CHECK-NEXT: $p7 = IMPLICIT_DEF - ; CHECK-NEXT: $p8 = IMPLICIT_DEF - ; CHECK-NEXT: $p9 = IMPLICIT_DEF - ; CHECK-NEXT: $p10 = IMPLICIT_DEF - ; CHECK-NEXT: $p11 = IMPLICIT_DEF - ; CHECK-NEXT: $p12 = IMPLICIT_DEF - ; CHECK-NEXT: $p13 = IMPLICIT_DEF - ; CHECK-NEXT: $p14 = IMPLICIT_DEF - ; CHECK-NEXT: $p15 = IMPLICIT_DEF - ; - ; CHECK-NEXT: $p0 = FILL_PPR_FROM_ZPR_SLOT_PSEUDO %stack.0, 0 :: (load (s128) from %stack.0) - ; - ; CHECK-NEXT: FAKE_USE implicit $z16, implicit $z17, implicit $z18, implicit $z19, implicit $z20, implicit $z21, implicit $z22, implicit $z23, implicit $z24, implicit $z25, implicit $z26, implicit $z27, implicit $z28, implicit $z29, implicit $z30, implicit $z31 - ; - ; CHECK-NEXT: RET_ReallyLR implicit $p0, implicit $z0, implicit $z1, implicit $z2, implicit $z3, implicit $z4, implicit $z5, implicit $z6, implicit $z7 - - ; EXPAND-LABEL: name: zpr_predicate_spill__spill_zpr - ; EXPAND: liveins: $p0, $z0, $z1, $z2, $z3, $z4, $z5, $z6, $z7, $fp, $p15, $p14, $p13, $p12, $p11, $p10, $p9, $p8, $p7, $p6, $p5, $p4, $z23, $z22, $z21, $z20, $z19, $z18, $z17, $z16 - ; EXPAND-NEXT: {{ $}} - ; - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1040, 0 - ; EXPAND-NEXT: frame-setup STRXui killed $fp, $sp, 128 :: (store (s64) into %stack.22) - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -20, implicit $vg - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p15, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 0 :: (store (s128) into %stack.21) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p14, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 1 :: (store (s128) into %stack.20) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p13, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 2 :: (store (s128) into %stack.19) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p12, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 3 :: (store (s128) into %stack.18) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p11, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 4 :: (store (s128) into %stack.17) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p10, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 5 :: (store (s128) into %stack.16) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p9, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 6 :: (store (s128) into %stack.15) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p8, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 7 :: (store (s128) into %stack.14) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p7, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 8 :: (store (s128) into %stack.13) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p6, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 9 :: (store (s128) into %stack.12) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p5, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 10 :: (store (s128) into %stack.11) - ; EXPAND-NEXT: $z24 = frame-setup CPY_ZPzI_B killed $p4, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z24, $sp, 11 :: (store (s128) into %stack.10) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z23, $sp, 12 :: (store (s128) into %stack.9) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z22, $sp, 13 :: (store (s128) into %stack.8) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z21, $sp, 14 :: (store (s128) into %stack.7) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z20, $sp, 15 :: (store (s128) into %stack.6) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z19, $sp, 16 :: (store (s128) into %stack.5) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z18, $sp, 17 :: (store (s128) into %stack.4) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z17, $sp, 18 :: (store (s128) into %stack.3) - ; EXPAND-NEXT: frame-setup STR_ZXI killed $z16, $sp, 19 :: (store (s128) into %stack.2) - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -2, implicit $vg - ; - ; EXPAND-NEXT: $z16 = IMPLICIT_DEF - ; EXPAND-NEXT: $z17 = IMPLICIT_DEF - ; EXPAND-NEXT: $z18 = IMPLICIT_DEF - ; EXPAND-NEXT: $z19 = IMPLICIT_DEF - ; EXPAND-NEXT: $z20 = IMPLICIT_DEF - ; EXPAND-NEXT: $z21 = IMPLICIT_DEF - ; EXPAND-NEXT: $z22 = IMPLICIT_DEF - ; EXPAND-NEXT: $z23 = IMPLICIT_DEF - ; EXPAND-NEXT: $z24 = IMPLICIT_DEF - ; EXPAND-NEXT: $z25 = IMPLICIT_DEF - ; EXPAND-NEXT: $z26 = IMPLICIT_DEF - ; EXPAND-NEXT: $z27 = IMPLICIT_DEF - ; EXPAND-NEXT: $z28 = IMPLICIT_DEF - ; EXPAND-NEXT: $z29 = IMPLICIT_DEF - ; EXPAND-NEXT: $z30 = IMPLICIT_DEF - ; EXPAND-NEXT: $z31 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $x8 = ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: STR_ZXI $z0, $x8, 0 :: (store (s128) into %stack.24) - ; EXPAND-NEXT: $z0 = CPY_ZPzI_B $p0, 1, 0 - ; EXPAND-NEXT: STR_ZXI $z0, $x8, 1 :: (store (s128) into %stack.0) - ; EXPAND-NEXT: $z0 = LDR_ZXI $x8, 0 :: (load (s128) from %stack.24) - ; - ; EXPAND-NEXT: $p0 = IMPLICIT_DEF - ; EXPAND-NEXT: $p1 = IMPLICIT_DEF - ; EXPAND-NEXT: $p2 = IMPLICIT_DEF - ; EXPAND-NEXT: $p3 = IMPLICIT_DEF - ; EXPAND-NEXT: $p4 = IMPLICIT_DEF - ; EXPAND-NEXT: $p5 = IMPLICIT_DEF - ; EXPAND-NEXT: $p6 = IMPLICIT_DEF - ; EXPAND-NEXT: $p7 = IMPLICIT_DEF - ; EXPAND-NEXT: $p8 = IMPLICIT_DEF - ; EXPAND-NEXT: $p9 = IMPLICIT_DEF - ; EXPAND-NEXT: $p10 = IMPLICIT_DEF - ; EXPAND-NEXT: $p11 = IMPLICIT_DEF - ; EXPAND-NEXT: $p12 = IMPLICIT_DEF - ; EXPAND-NEXT: $p13 = IMPLICIT_DEF - ; EXPAND-NEXT: $p14 = IMPLICIT_DEF - ; EXPAND-NEXT: $p15 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: STR_ZXI $z0, $x8, 0 :: (store (s128) into %stack.24) - ; EXPAND-NEXT: $z0 = LDR_ZXI $x8, 1 :: (load (s128) from %stack.0) - ; EXPAND-NEXT: $p0 = PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p0 = CMPNE_PPzZI_B $p0, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = LDR_ZXI killed $x8, 0 :: (load (s128) from %stack.24) - ; - ; EXPAND-NEXT: FAKE_USE implicit $z16, implicit $z17, implicit $z18, implicit $z19, implicit $z20, implicit $z21, implicit $z22, implicit $z23, implicit $z24, implicit $z25, implicit $z26, implicit $z27, implicit $z28, implicit $z29, implicit $z30, implicit $z31 - ; - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 2, implicit $vg - ; EXPAND-NEXT: $z23 = frame-destroy LDR_ZXI $sp, 12 :: (load (s128) from %stack.9) - ; EXPAND-NEXT: $z22 = frame-destroy LDR_ZXI $sp, 13 :: (load (s128) from %stack.8) - ; EXPAND-NEXT: $z21 = frame-destroy LDR_ZXI $sp, 14 :: (load (s128) from %stack.7) - ; EXPAND-NEXT: $z20 = frame-destroy LDR_ZXI $sp, 15 :: (load (s128) from %stack.6) - ; EXPAND-NEXT: $z19 = frame-destroy LDR_ZXI $sp, 16 :: (load (s128) from %stack.5) - ; EXPAND-NEXT: $z18 = frame-destroy LDR_ZXI $sp, 17 :: (load (s128) from %stack.4) - ; EXPAND-NEXT: $z17 = frame-destroy LDR_ZXI $sp, 18 :: (load (s128) from %stack.3) - ; EXPAND-NEXT: $z16 = frame-destroy LDR_ZXI $sp, 19 :: (load (s128) from %stack.2) - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 0 :: (load (s128) from %stack.21) - ; EXPAND-NEXT: $p1 = frame-destroy PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p15 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 1 :: (load (s128) from %stack.20) - ; EXPAND-NEXT: $p14 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 2 :: (load (s128) from %stack.19) - ; EXPAND-NEXT: $p13 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 3 :: (load (s128) from %stack.18) - ; EXPAND-NEXT: $p12 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 4 :: (load (s128) from %stack.17) - ; EXPAND-NEXT: $p11 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 5 :: (load (s128) from %stack.16) - ; EXPAND-NEXT: $p10 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 6 :: (load (s128) from %stack.15) - ; EXPAND-NEXT: $p9 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 7 :: (load (s128) from %stack.14) - ; EXPAND-NEXT: $p8 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 8 :: (load (s128) from %stack.13) - ; EXPAND-NEXT: $p7 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 9 :: (load (s128) from %stack.12) - ; EXPAND-NEXT: $p6 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 10 :: (load (s128) from %stack.11) - ; EXPAND-NEXT: $p5 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z24 = frame-destroy LDR_ZXI $sp, 11 :: (load (s128) from %stack.10) - ; EXPAND-NEXT: $p4 = frame-destroy CMPNE_PPzZI_B $p1, $z24, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 20, implicit $vg - ; EXPAND-NEXT: $fp = frame-destroy LDRXui $sp, 128 :: (load (s64) from %stack.22) - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1040, 0 - ; EXPAND-NEXT: RET undef $lr, implicit $p0, implicit $z0, implicit $z1, implicit $z2, implicit $z3, implicit $z4, implicit $z5, implicit $z6, implicit $z7 - $z16 = IMPLICIT_DEF - $z17 = IMPLICIT_DEF - $z18 = IMPLICIT_DEF - $z19 = IMPLICIT_DEF - $z20 = IMPLICIT_DEF - $z21 = IMPLICIT_DEF - $z22 = IMPLICIT_DEF - $z23 = IMPLICIT_DEF - $z24 = IMPLICIT_DEF - $z25 = IMPLICIT_DEF - $z26 = IMPLICIT_DEF - $z27 = IMPLICIT_DEF - $z28 = IMPLICIT_DEF - $z29 = IMPLICIT_DEF - $z30 = IMPLICIT_DEF - $z31 = IMPLICIT_DEF - - %1:ppr = COPY $p0 - - $p0 = IMPLICIT_DEF - $p1 = IMPLICIT_DEF - $p2 = IMPLICIT_DEF - $p3 = IMPLICIT_DEF - $p4 = IMPLICIT_DEF - $p5 = IMPLICIT_DEF - $p6 = IMPLICIT_DEF - $p7 = IMPLICIT_DEF - $p8 = IMPLICIT_DEF - $p9 = IMPLICIT_DEF - $p10 = IMPLICIT_DEF - $p11 = IMPLICIT_DEF - $p12 = IMPLICIT_DEF - $p13 = IMPLICIT_DEF - $p14 = IMPLICIT_DEF - $p15 = IMPLICIT_DEF - - $p0 = COPY %1 - - FAKE_USE implicit $z16, implicit $z17, implicit $z18, implicit $z19, implicit $z20, implicit $z21, implicit $z22, implicit $z23, implicit $z24, implicit $z25, implicit $z26, implicit $z27, implicit $z28, implicit $z29, implicit $z30, implicit $z31 - - RET_ReallyLR implicit $p0, implicit $z0, implicit $z1, implicit $z2, implicit $z3, implicit $z4, implicit $z5, implicit $z6, implicit $z7 -... ---- -name: zpr_predicate_spill_above_p7 -tracksRegLiveness: true -stack: -liveins: - - { reg: '$p0' } - - { reg: '$p1' } - - { reg: '$p2' } - - { reg: '$p3' } -body: | - bb.0.entry: - liveins: $p0, $p1, $p2, $p3 - - ; CHECK-LABEL: name: zpr_predicate_spill_above_p7 - ; CHECK: stack: - ; CHECK: - { id: 0, name: '', type: spill-slot, offset: 0, size: 16, alignment: 16, - ; CHECK-NEXT: stack-id: scalable-vector, callee-saved-register: - ; CHECK: liveins: $p0, $p1, $p2, $p3 - ; CHECK-NEXT: {{ $}} - ; - ; CHECK-NEXT: $p15 = IMPLICIT_DEF - ; - ; CHECK-NEXT: SPILL_PPR_TO_ZPR_SLOT_PSEUDO $p15, %stack.0, 0 :: (store (s128) into %stack.0) - ; - ; CHECK-NEXT: $p0 = IMPLICIT_DEF - ; CHECK-NEXT: $p1 = IMPLICIT_DEF - ; CHECK-NEXT: $p2 = IMPLICIT_DEF - ; CHECK-NEXT: $p3 = IMPLICIT_DEF - ; CHECK-NEXT: $p4 = IMPLICIT_DEF - ; CHECK-NEXT: $p5 = IMPLICIT_DEF - ; CHECK-NEXT: $p6 = IMPLICIT_DEF - ; CHECK-NEXT: $p7 = IMPLICIT_DEF - ; CHECK-NEXT: $p8 = IMPLICIT_DEF - ; CHECK-NEXT: $p9 = IMPLICIT_DEF - ; CHECK-NEXT: $p10 = IMPLICIT_DEF - ; CHECK-NEXT: $p11 = IMPLICIT_DEF - ; CHECK-NEXT: $p12 = IMPLICIT_DEF - ; CHECK-NEXT: $p13 = IMPLICIT_DEF - ; CHECK-NEXT: $p14 = IMPLICIT_DEF - ; CHECK-NEXT: $p15 = IMPLICIT_DEF - ; - ; CHECK-NEXT: $p15 = FILL_PPR_FROM_ZPR_SLOT_PSEUDO %stack.0, 0 :: (load (s128) from %stack.0) - ; - ; CHECK-NEXT: FAKE_USE implicit $p4, implicit $p5, implicit $p6, implicit $p7 - ; - ; CHECK-NEXT: RET_ReallyLR implicit $p0, implicit $p1, implicit $p2, implicit $p3 - - ; EXPAND-LABEL: name: zpr_predicate_spill_above_p7 - ; EXPAND: liveins: $p0, $p1, $p2, $p3, $fp, $p15, $p14, $p13, $p12, $p11, $p10, $p9, $p8, $p7, $p6, $p5, $p4 - ; EXPAND-NEXT: {{ $}} - ; - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1040, 0 - ; EXPAND-NEXT: frame-setup STRXui killed $fp, $sp, 128 :: (store (s64) into %stack.14) - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -12, implicit $vg - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p15, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 0 :: (store (s128) into %stack.13) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p14, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 1 :: (store (s128) into %stack.12) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p13, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 2 :: (store (s128) into %stack.11) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p12, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 3 :: (store (s128) into %stack.10) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p11, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 4 :: (store (s128) into %stack.9) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p10, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 5 :: (store (s128) into %stack.8) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p9, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 6 :: (store (s128) into %stack.7) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p8, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 7 :: (store (s128) into %stack.6) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p7, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 8 :: (store (s128) into %stack.5) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p6, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 9 :: (store (s128) into %stack.4) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p5, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 10 :: (store (s128) into %stack.3) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p4, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 11 :: (store (s128) into %stack.2) - ; EXPAND-NEXT: $sp = frame-setup SUBXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -2, implicit $vg - ; - ; EXPAND-NEXT: $p15 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = CPY_ZPzI_B $p15, 1, 0 - ; EXPAND-NEXT: $x8 = ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: STR_ZXI $z0, $x8, 1 :: (store (s128) into %stack.0) - ; - ; EXPAND-NEXT: $p0 = IMPLICIT_DEF - ; EXPAND-NEXT: $p1 = IMPLICIT_DEF - ; EXPAND-NEXT: $p2 = IMPLICIT_DEF - ; EXPAND-NEXT: $p3 = IMPLICIT_DEF - ; EXPAND-NEXT: $p4 = IMPLICIT_DEF - ; EXPAND-NEXT: $p5 = IMPLICIT_DEF - ; EXPAND-NEXT: $p6 = IMPLICIT_DEF - ; EXPAND-NEXT: $p7 = IMPLICIT_DEF - ; EXPAND-NEXT: $p8 = IMPLICIT_DEF - ; EXPAND-NEXT: $p9 = IMPLICIT_DEF - ; EXPAND-NEXT: $p10 = IMPLICIT_DEF - ; EXPAND-NEXT: $p11 = IMPLICIT_DEF - ; EXPAND-NEXT: $p12 = IMPLICIT_DEF - ; EXPAND-NEXT: $p13 = IMPLICIT_DEF - ; EXPAND-NEXT: $p14 = IMPLICIT_DEF - ; EXPAND-NEXT: $p15 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = CPY_ZPzI_B $p0, 1, 0 - ; EXPAND-NEXT: STR_ZXI $z0, $x8, 0 :: (store (s128) into %stack.16) - ; EXPAND-NEXT: $z0 = LDR_ZXI $x8, 1 :: (load (s128) from %stack.0) - ; EXPAND-NEXT: $p0 = PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p15 = CMPNE_PPzZI_B $p0, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = LDR_ZXI killed $x8, 0 :: (load (s128) from %stack.16) - ; EXPAND-NEXT: $p0 = PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p0 = CMPNE_PPzZI_B $p0, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; - ; EXPAND-NEXT: FAKE_USE implicit $p4, implicit $p5, implicit $p6, implicit $p7 - ; - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1024, 0 - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 2, implicit $vg - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 0 :: (load (s128) from %stack.13) - ; EXPAND-NEXT: $p4 = frame-destroy PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p15 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 1 :: (load (s128) from %stack.12) - ; EXPAND-NEXT: $p14 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 2 :: (load (s128) from %stack.11) - ; EXPAND-NEXT: $p13 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 3 :: (load (s128) from %stack.10) - ; EXPAND-NEXT: $p12 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 4 :: (load (s128) from %stack.9) - ; EXPAND-NEXT: $p11 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 5 :: (load (s128) from %stack.8) - ; EXPAND-NEXT: $p10 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 6 :: (load (s128) from %stack.7) - ; EXPAND-NEXT: $p9 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 7 :: (load (s128) from %stack.6) - ; EXPAND-NEXT: $p8 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 8 :: (load (s128) from %stack.5) - ; EXPAND-NEXT: $p7 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 9 :: (load (s128) from %stack.4) - ; EXPAND-NEXT: $p6 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 10 :: (load (s128) from %stack.3) - ; EXPAND-NEXT: $p5 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 11 :: (load (s128) from %stack.2) - ; EXPAND-NEXT: $p4 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 12, implicit $vg - ; EXPAND-NEXT: $fp = frame-destroy LDRXui $sp, 128 :: (load (s64) from %stack.14) - ; EXPAND-NEXT: $sp = frame-destroy ADDXri $sp, 1040, 0 - ; EXPAND-NEXT: RET undef $lr, implicit $p0, implicit $p1, implicit $p2, implicit $p3 - $p15 = IMPLICIT_DEF - %1:ppr = COPY $p15 - - $p0 = IMPLICIT_DEF - $p1 = IMPLICIT_DEF - $p2 = IMPLICIT_DEF - $p3 = IMPLICIT_DEF - $p4 = IMPLICIT_DEF - $p5 = IMPLICIT_DEF - $p6 = IMPLICIT_DEF - $p7 = IMPLICIT_DEF - $p8 = IMPLICIT_DEF - $p9 = IMPLICIT_DEF - $p10 = IMPLICIT_DEF - $p11 = IMPLICIT_DEF - $p12 = IMPLICIT_DEF - $p13 = IMPLICIT_DEF - $p14 = IMPLICIT_DEF - $p15 = IMPLICIT_DEF - - $p15 = COPY %1 - - FAKE_USE implicit $p4, implicit $p5, implicit $p6, implicit $p7 - - RET_ReallyLR implicit $p0, implicit $p1, implicit $p2, implicit $p3 -... ---- -name: zpr_predicate_spill_p4_saved -tracksRegLiveness: true -stack: -liveins: - - { reg: '$p0' } - - { reg: '$p1' } - - { reg: '$p2' } - - { reg: '$p3' } -body: | - bb.0.entry: - liveins: $p0, $p1, $p2, $p3 - - ; CHECK-LABEL: name: zpr_predicate_spill_p4_saved - ; CHECK: liveins: $p0, $p1, $p2, $p3 - ; CHECK-NEXT: {{ $}} - ; - ; CHECK-NEXT: $p8 = IMPLICIT_DEF - ; - ; CHECK-NEXT: RET_ReallyLR implicit $p0, implicit $p1, implicit $p2, implicit $p3 - - ; EXPAND-LABEL: name: zpr_predicate_spill_p4_saved - ; EXPAND: liveins: $p0, $p1, $p2, $p3, $fp, $p8, $p4 - ; EXPAND-NEXT: {{ $}} - ; EXPAND-NEXT: early-clobber $sp = frame-setup STRXpre killed $fp, $sp, -16 :: (store (s64) into %stack.2) - ; EXPAND-NEXT: $sp = frame-setup ADDVL_XXI $sp, -2, implicit $vg - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p8, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 0 :: (store (s128) into %stack.1) - ; EXPAND-NEXT: $z0 = frame-setup CPY_ZPzI_B killed $p4, 1, 0 - ; EXPAND-NEXT: frame-setup STR_ZXI $z0, $sp, 1 :: (store (s128) into %stack.0) - ; - ; EXPAND-NEXT: $p8 = IMPLICIT_DEF - ; - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 0 :: (load (s128) from %stack.1) - ; EXPAND-NEXT: $p4 = frame-destroy PTRUE_B 31, implicit $vg - ; EXPAND-NEXT: $p8 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $z0 = frame-destroy LDR_ZXI $sp, 1 :: (load (s128) from %stack.0) - ; EXPAND-NEXT: $p4 = frame-destroy CMPNE_PPzZI_B $p4, $z0, 0, implicit-def $nzcv, implicit-def $nzcv - ; EXPAND-NEXT: $sp = frame-destroy ADDVL_XXI $sp, 2, implicit $vg - ; EXPAND-NEXT: early-clobber $sp, $fp = frame-destroy LDRXpost $sp, 16 :: (load (s64) from %stack.2) - ; EXPAND-NEXT: RET undef $lr, implicit $p0, implicit $p1, implicit $p2, implicit $p3 - - ; If we spill a register above p8, p4 must also be saved, so we can guarantee - ; they will be a register (in the range p0-p7 to for the cmpne reload). - $p8 = IMPLICIT_DEF - - RET_ReallyLR implicit $p0, implicit $p1, implicit $p2, implicit $p3 -... diff --git a/llvm/test/CodeGen/AArch64/ssve-stack-hazard-remarks.ll b/llvm/test/CodeGen/AArch64/ssve-stack-hazard-remarks.ll index 01e3d3a..c0a2943 100644 --- a/llvm/test/CodeGen/AArch64/ssve-stack-hazard-remarks.ll +++ b/llvm/test/CodeGen/AArch64/ssve-stack-hazard-remarks.ll @@ -1,7 +1,5 @@ ; RUN: llc < %s -mtriple=aarch64 -mattr=+sve2 -pass-remarks-analysis=sme -aarch64-stack-hazard-remark-size=64 -o /dev/null < %s 2>&1 | FileCheck %s --check-prefixes=CHECK ; RUN: llc < %s -mtriple=aarch64 -mattr=+sve2 -pass-remarks-analysis=sme -aarch64-stack-hazard-size=1024 -o /dev/null < %s 2>&1 | FileCheck %s --check-prefixes=CHECK-PADDING -; RUN: llc < %s -mtriple=aarch64 -mattr=+sve2 -pass-remarks-analysis=sme -aarch64-enable-zpr-predicate-spills -aarch64-stack-hazard-remark-size=64 -o /dev/null < %s 2>&1 | FileCheck %s --check-prefixes=CHECK-ZPR-PRED-SPILLS -; RUN: llc < %s -mtriple=aarch64 -mattr=+sve2 -pass-remarks-analysis=sme -aarch64-enable-zpr-predicate-spills -aarch64-stack-hazard-size=1024 -o /dev/null < %s 2>&1 | FileCheck %s --check-prefixes=CHECK-ZPR-PRED-SPILLS-WITH-PADDING ; Don't emit remarks for non-streaming functions. define float @csr_x20_stackargs_notsc(float %a, float %b, float %c, float %d, float %e, float %f, float %g, float %h, float %i) { @@ -69,16 +67,11 @@ entry: ; SVE calling conventions ; Padding is placed between predicate and fpr/zpr register spills, so only emit remarks when hazard padding is off. -; Note: The -aarch64-enable-zpr-predicate-spills option is deprecated (and will be removed soon). define i32 @svecc_call(<4 x i16> %P0, ptr %P1, i32 %P2, <vscale x 16 x i8> %P3, i16 %P4) #2 { ; CHECK: remark: <unknown>:0:0: stack hazard in 'svecc_call': PPR stack object at [SP-64-258 * vscale] is too close to FPR stack object at [SP-64-256 * vscale] ; CHECK: remark: <unknown>:0:0: stack hazard in 'svecc_call': FPR stack object at [SP-64-16 * vscale] is too close to GPR stack object at [SP-64] ; CHECK-PADDING-NOT: remark: <unknown>:0:0: stack hazard in 'svecc_call': -; CHECK-ZPR-PRED-SPILLS-NOT: <unknown>:0:0: stack hazard in 'svecc_call': PPR stack object at {{.*}} is too close to FPR stack object -; CHECK-ZPR-PRED-SPILLS: <unknown>:0:0: stack hazard in 'svecc_call': FPR stack object at [SP-64-16 * vscale] is too close to GPR stack object at [SP-64] -; CHECK-ZPR-PRED-SPILLS-WITH-PADDING-NOT: <unknown>:0:0: stack hazard in 'svecc_call': PPR stack object at {{.*}} is too close to FPR stack object -; CHECK-ZPR-PRED-SPILLS-WITH-PADDING-NOT: <unknown>:0:0: stack hazard in 'svecc_call': FPR stack object at {{.*}} is too close to GPR stack object entry: tail call void asm sideeffect "", "~{x0},~{x28},~{x27},~{x3}"() #2 %call = call ptr @memset(ptr noundef nonnull %P1, i32 noundef 45, i32 noundef 37) @@ -89,10 +82,6 @@ define i32 @svecc_alloca_call(<4 x i16> %P0, ptr %P1, i32 %P2, <vscale x 16 x i8 ; CHECK: remark: <unknown>:0:0: stack hazard in 'svecc_alloca_call': PPR stack object at [SP-64-258 * vscale] is too close to FPR stack object at [SP-64-256 * vscale] ; CHECK: remark: <unknown>:0:0: stack hazard in 'svecc_alloca_call': FPR stack object at [SP-64-16 * vscale] is too close to GPR stack object at [SP-64] ; CHECK-PADDING-NOT: remark: <unknown>:0:0: stack hazard in 'svecc_alloca_call': -; CHECK-ZPR-PRED-SPILLS-NOT: <unknown>:0:0: stack hazard in 'svecc_call': PPR stack object at {{.*}} is too close to FPR stack object -; CHECK-ZPR-PRED-SPILLS: <unknown>:0:0: stack hazard in 'svecc_alloca_call': FPR stack object at [SP-64-16 * vscale] is too close to GPR stack object at [SP-64] -; CHECK-ZPR-PRED-SPILLS-WITH-PADDING-NOT: <unknown>:0:0: stack hazard in 'svecc_alloca_call': PPR stack object at {{.*}} is too close to FPR stack object -; CHECK-ZPR-PRED-SPILLS-WITH-PADDING-NOT: <unknown>:0:0: stack hazard in 'svecc_alloca_call': FPR stack object at {{.*}} is too close to GPR stack object entry: tail call void asm sideeffect "", "~{x0},~{x28},~{x27},~{x3}"() #2 %0 = alloca [37 x i8], align 16 diff --git a/llvm/test/CodeGen/AMDGPU/agpr-copy-no-free-registers.ll b/llvm/test/CodeGen/AMDGPU/agpr-copy-no-free-registers.ll index 9e24023..ebbeab9 100644 --- a/llvm/test/CodeGen/AMDGPU/agpr-copy-no-free-registers.ll +++ b/llvm/test/CodeGen/AMDGPU/agpr-copy-no-free-registers.ll @@ -146,9 +146,9 @@ define void @no_free_vgprs_at_agpr_to_agpr_copy(float %v0, float %v1) #0 { ; GFX908-NEXT: ;;#ASMSTART ; GFX908-NEXT: ; copy ; GFX908-NEXT: ;;#ASMEND -; GFX908-NEXT: v_accvgpr_read_b32 v32, a2 +; GFX908-NEXT: v_accvgpr_read_b32 v39, a2 ; GFX908-NEXT: s_nop 1 -; GFX908-NEXT: v_accvgpr_write_b32 a3, v32 +; GFX908-NEXT: v_accvgpr_write_b32 a3, v39 ; GFX908-NEXT: ;;#ASMSTART ; GFX908-NEXT: ; use a3 v[0:31] ; GFX908-NEXT: ;;#ASMEND @@ -437,9 +437,9 @@ define void @v32_asm_def_use(float %v0, float %v1) #4 { ; GFX908-NEXT: ; copy ; GFX908-NEXT: ;;#ASMEND ; GFX908-NEXT: s_nop 7 -; GFX908-NEXT: v_accvgpr_read_b32 v33, a2 +; GFX908-NEXT: v_accvgpr_read_b32 v35, a2 ; GFX908-NEXT: s_nop 1 -; GFX908-NEXT: v_accvgpr_write_b32 a3, v33 +; GFX908-NEXT: v_accvgpr_write_b32 a3, v35 ; GFX908-NEXT: ;;#ASMSTART ; GFX908-NEXT: ; use a3 v[0:31] ; GFX908-NEXT: ;;#ASMEND @@ -1045,9 +1045,9 @@ define void @no_free_vgprs_at_sgpr_to_agpr_copy(float %v0, float %v1) #0 { ; GFX908-NEXT: ;;#ASMSTART ; GFX908-NEXT: ; copy ; GFX908-NEXT: ;;#ASMEND -; GFX908-NEXT: v_accvgpr_read_b32 v32, a2 +; GFX908-NEXT: v_accvgpr_read_b32 v39, a2 ; GFX908-NEXT: s_nop 1 -; GFX908-NEXT: v_accvgpr_write_b32 a3, v32 +; GFX908-NEXT: v_accvgpr_write_b32 a3, v39 ; GFX908-NEXT: ;;#ASMSTART ; GFX908-NEXT: ; use a3 v[0:31] ; GFX908-NEXT: ;;#ASMEND diff --git a/llvm/test/CodeGen/AMDGPU/agpr-copy-propagation.mir b/llvm/test/CodeGen/AMDGPU/agpr-copy-propagation.mir index a42cf43..7e82382d 100644 --- a/llvm/test/CodeGen/AMDGPU/agpr-copy-propagation.mir +++ b/llvm/test/CodeGen/AMDGPU/agpr-copy-propagation.mir @@ -40,8 +40,8 @@ body: | ; GFX908: liveins: $agpr0 ; GFX908-NEXT: {{ $}} ; GFX908-NEXT: renamable $vgpr0 = COPY renamable $agpr0, implicit $exec - ; GFX908-NEXT: renamable $agpr1 = COPY renamable $vgpr0, implicit $exec - ; GFX908-NEXT: renamable $agpr2 = COPY renamable $vgpr0, implicit $exec + ; GFX908-NEXT: renamable $agpr1 = COPY $agpr0, implicit $exec + ; GFX908-NEXT: renamable $agpr2 = COPY $agpr0, implicit $exec ; GFX908-NEXT: S_ENDPGM 0, implicit $vgpr0, implicit $agpr1, implicit $agpr2 ; ; GFX90A-LABEL: name: do_not_propagate_agpr_to_agpr diff --git a/llvm/test/CodeGen/AMDGPU/mfma-no-register-aliasing.ll b/llvm/test/CodeGen/AMDGPU/mfma-no-register-aliasing.ll index 51cd564..f46116e 100644 --- a/llvm/test/CodeGen/AMDGPU/mfma-no-register-aliasing.ll +++ b/llvm/test/CodeGen/AMDGPU/mfma-no-register-aliasing.ll @@ -95,66 +95,66 @@ define amdgpu_kernel void @test_mfma_f32_32x32x1f32(ptr addrspace(1) %arg) #0 { ; GREEDY908-NEXT: v_mfma_f32_32x32x1f32 a[32:63], v3, v0, a[0:31] ; GREEDY908-NEXT: s_nop 15 ; GREEDY908-NEXT: s_nop 1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a32 -; GREEDY908-NEXT: v_accvgpr_read_b32 v5, a61 -; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a60 -; GREEDY908-NEXT: v_accvgpr_write_b32 a2, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a33 -; GREEDY908-NEXT: v_accvgpr_read_b32 v7, a59 -; GREEDY908-NEXT: v_accvgpr_read_b32 v8, a58 -; GREEDY908-NEXT: v_accvgpr_write_b32 a3, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a32 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a33 ; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a34 -; GREEDY908-NEXT: v_accvgpr_read_b32 v9, a57 -; GREEDY908-NEXT: v_accvgpr_read_b32 v10, a56 +; GREEDY908-NEXT: v_accvgpr_write_b32 a2, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a3, v6 ; GREEDY908-NEXT: v_accvgpr_write_b32 a4, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a35 -; GREEDY908-NEXT: v_accvgpr_read_b32 v11, a55 -; GREEDY908-NEXT: v_accvgpr_read_b32 v12, a54 -; GREEDY908-NEXT: v_accvgpr_write_b32 a5, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a36 -; GREEDY908-NEXT: v_accvgpr_read_b32 v13, a53 -; GREEDY908-NEXT: v_accvgpr_read_b32 v14, a52 -; GREEDY908-NEXT: v_accvgpr_write_b32 a6, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a35 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a36 ; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a37 -; GREEDY908-NEXT: v_accvgpr_read_b32 v15, a51 -; GREEDY908-NEXT: v_accvgpr_read_b32 v16, a50 +; GREEDY908-NEXT: v_accvgpr_write_b32 a5, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a6, v6 ; GREEDY908-NEXT: v_accvgpr_write_b32 a7, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a38 -; GREEDY908-NEXT: v_accvgpr_read_b32 v17, a49 -; GREEDY908-NEXT: v_accvgpr_read_b32 v18, a48 -; GREEDY908-NEXT: v_accvgpr_write_b32 a8, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a39 -; GREEDY908-NEXT: v_accvgpr_read_b32 v19, a47 -; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a46 -; GREEDY908-NEXT: v_accvgpr_write_b32 a9, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a38 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a39 ; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a40 -; GREEDY908-NEXT: v_accvgpr_write_b32 a16, v2 -; GREEDY908-NEXT: v_accvgpr_write_b32 a17, v19 +; GREEDY908-NEXT: v_accvgpr_write_b32 a8, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a9, v6 ; GREEDY908-NEXT: v_accvgpr_write_b32 a10, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a41 -; GREEDY908-NEXT: v_accvgpr_write_b32 a18, v18 -; GREEDY908-NEXT: v_accvgpr_write_b32 a19, v17 -; GREEDY908-NEXT: v_accvgpr_write_b32 a11, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a42 -; GREEDY908-NEXT: v_accvgpr_write_b32 a20, v16 -; GREEDY908-NEXT: v_accvgpr_write_b32 a21, v15 -; GREEDY908-NEXT: v_accvgpr_write_b32 a12, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a41 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a42 ; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a43 -; GREEDY908-NEXT: v_accvgpr_write_b32 a22, v14 -; GREEDY908-NEXT: v_accvgpr_write_b32 a23, v13 +; GREEDY908-NEXT: v_accvgpr_write_b32 a11, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a12, v6 ; GREEDY908-NEXT: v_accvgpr_write_b32 a13, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a44 -; GREEDY908-NEXT: v_accvgpr_write_b32 a24, v12 -; GREEDY908-NEXT: v_accvgpr_write_b32 a25, v11 -; GREEDY908-NEXT: v_accvgpr_write_b32 a14, v1 -; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a45 -; GREEDY908-NEXT: v_accvgpr_write_b32 a26, v10 -; GREEDY908-NEXT: v_accvgpr_write_b32 a27, v9 -; GREEDY908-NEXT: v_accvgpr_write_b32 a15, v1 -; GREEDY908-NEXT: v_accvgpr_write_b32 a28, v8 -; GREEDY908-NEXT: v_accvgpr_write_b32 a29, v7 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a44 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a45 +; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a46 +; GREEDY908-NEXT: v_accvgpr_write_b32 a14, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a15, v6 +; GREEDY908-NEXT: v_accvgpr_write_b32 a16, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a47 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a48 +; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a49 +; GREEDY908-NEXT: v_accvgpr_write_b32 a17, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a18, v6 +; GREEDY908-NEXT: v_accvgpr_write_b32 a19, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a50 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a51 +; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a52 +; GREEDY908-NEXT: v_accvgpr_write_b32 a20, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a21, v6 +; GREEDY908-NEXT: v_accvgpr_write_b32 a22, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a53 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a54 +; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a55 +; GREEDY908-NEXT: v_accvgpr_write_b32 a23, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a24, v6 +; GREEDY908-NEXT: v_accvgpr_write_b32 a25, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a56 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a57 +; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a58 +; GREEDY908-NEXT: v_accvgpr_write_b32 a26, v2 +; GREEDY908-NEXT: v_accvgpr_write_b32 a27, v6 +; GREEDY908-NEXT: v_accvgpr_write_b32 a28, v1 +; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a59 +; GREEDY908-NEXT: v_accvgpr_read_b32 v6, a60 +; GREEDY908-NEXT: v_accvgpr_read_b32 v1, a61 +; GREEDY908-NEXT: v_accvgpr_write_b32 a29, v2 ; GREEDY908-NEXT: v_accvgpr_write_b32 a30, v6 -; GREEDY908-NEXT: v_accvgpr_write_b32 a31, v5 +; GREEDY908-NEXT: v_accvgpr_write_b32 a31, v1 ; GREEDY908-NEXT: s_nop 0 ; GREEDY908-NEXT: v_mfma_f32_32x32x1f32 a[0:31], v3, v0, a[0:31] ; GREEDY908-NEXT: s_nop 15 @@ -667,11 +667,11 @@ define amdgpu_kernel void @test_mfma_f32_16x16x1f32(ptr addrspace(1) %arg) #0 { ; GREEDY908-NEXT: v_mfma_f32_16x16x1f32 a[18:33], v0, v1, a[18:33] ; GREEDY908-NEXT: v_mfma_f32_16x16x1f32 a[2:17], v0, v1, a[18:33] ; GREEDY908-NEXT: s_nop 8 +; GREEDY908-NEXT: v_accvgpr_read_b32 v5, a18 ; GREEDY908-NEXT: v_accvgpr_read_b32 v2, a19 -; GREEDY908-NEXT: v_accvgpr_read_b32 v3, a18 ; GREEDY908-NEXT: s_nop 0 +; GREEDY908-NEXT: v_accvgpr_write_b32 a0, v5 ; GREEDY908-NEXT: v_accvgpr_write_b32 a1, v2 -; GREEDY908-NEXT: v_accvgpr_write_b32 a0, v3 ; GREEDY908-NEXT: s_nop 0 ; GREEDY908-NEXT: v_mfma_f32_16x16x1f32 a[0:15], v0, v1, a[0:15] ; GREEDY908-NEXT: s_nop 9 diff --git a/llvm/test/CodeGen/AMDGPU/no-fold-accvgpr-mov.ll b/llvm/test/CodeGen/AMDGPU/no-fold-accvgpr-mov.ll index cf244f0..be1788c 100644 --- a/llvm/test/CodeGen/AMDGPU/no-fold-accvgpr-mov.ll +++ b/llvm/test/CodeGen/AMDGPU/no-fold-accvgpr-mov.ll @@ -54,19 +54,20 @@ define amdgpu_kernel void @matmul_kernel(i32 %a0, i32 %a1) { ; GFX908-NEXT: s_branch .LBB0_2 ; GFX908-NEXT: .LBB0_1: ; %bb2 ; GFX908-NEXT: ; in Loop: Header=BB0_2 Depth=1 +; GFX908-NEXT: s_nop 6 +; GFX908-NEXT: v_accvgpr_read_b32 v3, a2 ; GFX908-NEXT: s_or_b32 s4, s3, 1 ; GFX908-NEXT: s_ashr_i32 s5, s3, 31 ; GFX908-NEXT: s_mov_b32 s3, s2 ; GFX908-NEXT: v_mov_b32_e32 v1, s2 -; GFX908-NEXT: s_nop 2 -; GFX908-NEXT: v_accvgpr_read_b32 v0, a2 ; GFX908-NEXT: v_mov_b32_e32 v2, s3 +; GFX908-NEXT: v_accvgpr_write_b32 a0, v3 ; GFX908-NEXT: v_accvgpr_read_b32 v4, a1 ; GFX908-NEXT: v_accvgpr_read_b32 v3, a1 -; GFX908-NEXT: v_accvgpr_write_b32 a0, v0 +; GFX908-NEXT: s_and_b32 s3, s5, s4 ; GFX908-NEXT: v_accvgpr_write_b32 a2, v4 ; GFX908-NEXT: v_accvgpr_write_b32 a3, v3 -; GFX908-NEXT: s_and_b32 s3, s5, s4 +; GFX908-NEXT: s_nop 0 ; GFX908-NEXT: v_mfma_f32_16x16x16f16 a[2:5], v[1:2], v[1:2], a[0:3] ; GFX908-NEXT: s_cbranch_execz .LBB0_4 ; GFX908-NEXT: .LBB0_2: ; %bb diff --git a/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0-callable.ll b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0-callable.ll index 6b7d704..ede470b 100644 --- a/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0-callable.ll +++ b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0-callable.ll @@ -1,13 +1,11 @@ ; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1100 < %s | FileCheck --check-prefixes=CHECK,GFX11 %s ; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 < %s | FileCheck --check-prefixes=CHECK,GFX12 %s -; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 -mattr=+dynamic-vgpr < %s | FileCheck --check-prefixes=CHECK,GFX12,DVGPR %s ; CHECK: .amdgpu_pal_metadata ; CHECK-NEXT: --- ; CHECK-NEXT: amdpal.pipelines: ; CHECK-NEXT: - .api: Vulkan ; CHECK-NEXT: .compute_registers: -; DVGPR-NEXT: .dynamic_vgpr_en: true ; CHECK-NEXT: .tg_size_en: true ; CHECK-NEXT: .tgid_x_en: false ; CHECK-NEXT: .tgid_y_en: false diff --git a/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0.ll b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0.ll index 5c0c366..5325499 100644 --- a/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0.ll +++ b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.0.ll @@ -1,17 +1,14 @@ -; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1100 <%s | FileCheck %s --check-prefixes=CHECK,GFX11,NODVGPR -; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 <%s | FileCheck %s --check-prefixes=CHECK,NODVGPR -; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 -mattr=+dynamic-vgpr <%s | FileCheck %s --check-prefixes=CHECK,DVGPR +; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1100 <%s | FileCheck %s --check-prefixes=CHECK,GFX11 +; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 <%s | FileCheck %s --check-prefixes=CHECK ; CHECK-LABEL: {{^}}_amdgpu_cs_main: -; NODVGPR: ; TotalNumSgprs: 4 -; DVGPR: ; TotalNumSgprs: 34 +; CHECK: ; TotalNumSgprs: 4 ; CHECK: ; NumVgprs: 2 ; CHECK: .amdgpu_pal_metadata ; CHECK-NEXT: --- ; CHECK-NEXT: amdpal.pipelines: ; CHECK-NEXT: - .api: Vulkan ; CHECK-NEXT: .compute_registers: -; DVGPR-NEXT: .dynamic_vgpr_en: true ; CHECK-NEXT: .tg_size_en: true ; CHECK-NEXT: .tgid_x_en: false ; CHECK-NEXT: .tgid_y_en: false @@ -57,7 +54,6 @@ ; CHECK-NEXT: .cs: ; CHECK-NEXT: .checksum_value: 0x9444d7d0 ; CHECK-NEXT: .debug_mode: false -; DVGPR-NEXT: .dynamic_vgpr_saved_count: 0x70 ; CHECK-NEXT: .entry_point: _amdgpu_cs_main ; CHECK-NEXT: .entry_point_symbol: _amdgpu_cs_main ; CHECK-NEXT: .excp_en: 0 @@ -69,8 +65,7 @@ ; CHECK-NEXT: .mem_ordered: true ; CHECK-NEXT: .scratch_en: false ; CHECK-NEXT: .scratch_memory_size: 0 -; NODVGPR-NEXT: .sgpr_count: 0x4 -; DVGPR-NEXT: .sgpr_count: 0x22 +; CHECK-NEXT: .sgpr_count: 0x4 ; CHECK-NEXT: .sgpr_limit: 0x6a ; CHECK-NEXT: .threadgroup_dimensions: ; CHECK-NEXT: - 0x1 diff --git a/llvm/test/CodeGen/AMDGPU/pal-metadata-3.6-dvgpr.ll b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.6-dvgpr.ll new file mode 100644 index 0000000..e598b0c --- /dev/null +++ b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.6-dvgpr.ll @@ -0,0 +1,204 @@ +; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 <%s | FileCheck %s --check-prefixes=CHECK + +; CHECK-LABEL: {{^}}_amdgpu_cs_main: +; CHECK: ; TotalNumSgprs: 34 +; CHECK: ; NumVgprs: 2 +; CHECK: .amdgpu_pal_metadata +; CHECK-NEXT: --- +; CHECK-NEXT: amdpal.pipelines: +; CHECK-NEXT: - .api: Vulkan +; CHECK-NEXT: .compute_registers: +; CHECK-NEXT: .dynamic_vgpr_en: true +; CHECK-NEXT: .tg_size_en: true +; CHECK-NEXT: .tgid_x_en: false +; CHECK-NEXT: .tgid_y_en: false +; CHECK-NEXT: .tgid_z_en: false +; CHECK-NEXT: .tidig_comp_cnt: 0x1 +; CHECK-NEXT: .graphics_registers: +; CHECK-NEXT: .ps_extra_lds_size: 0 +; CHECK-NEXT: .spi_ps_input_addr: +; CHECK-NEXT: .ancillary_ena: false +; CHECK-NEXT: .front_face_ena: true +; CHECK-NEXT: .line_stipple_tex_ena: false +; CHECK-NEXT: .linear_center_ena: true +; CHECK-NEXT: .linear_centroid_ena: true +; CHECK-NEXT: .linear_sample_ena: true +; CHECK-NEXT: .persp_center_ena: true +; CHECK-NEXT: .persp_centroid_ena: true +; CHECK-NEXT: .persp_pull_model_ena: false +; CHECK-NEXT: .persp_sample_ena: true +; CHECK-NEXT: .pos_fixed_pt_ena: true +; CHECK-NEXT: .pos_w_float_ena: false +; CHECK-NEXT: .pos_x_float_ena: false +; CHECK-NEXT: .pos_y_float_ena: false +; CHECK-NEXT: .pos_z_float_ena: false +; CHECK-NEXT: .sample_coverage_ena: false +; CHECK-NEXT: .spi_ps_input_ena: +; CHECK-NEXT: .ancillary_ena: false +; CHECK-NEXT: .front_face_ena: false +; CHECK-NEXT: .line_stipple_tex_ena: false +; CHECK-NEXT: .linear_center_ena: false +; CHECK-NEXT: .linear_centroid_ena: false +; CHECK-NEXT: .linear_sample_ena: false +; CHECK-NEXT: .persp_center_ena: false +; CHECK-NEXT: .persp_centroid_ena: false +; CHECK-NEXT: .persp_pull_model_ena: false +; CHECK-NEXT: .persp_sample_ena: true +; CHECK-NEXT: .pos_fixed_pt_ena: false +; CHECK-NEXT: .pos_w_float_ena: false +; CHECK-NEXT: .pos_x_float_ena: false +; CHECK-NEXT: .pos_y_float_ena: false +; CHECK-NEXT: .pos_z_float_ena: false +; CHECK-NEXT: .sample_coverage_ena: false +; CHECK-NEXT: .hardware_stages: +; CHECK-NEXT: .cs: +; CHECK-NEXT: .checksum_value: 0x9444d7d0 +; CHECK-NEXT: .debug_mode: false +; CHECK-NEXT: .dynamic_vgpr_saved_count: 0x70 +; CHECK-NOT: .entry_point: _amdgpu_cs_main +; CHECK-NEXT: .entry_point_symbol: _amdgpu_cs_main +; CHECK-NEXT: .excp_en: 0 +; CHECK-NEXT: .float_mode: 0xc0 +; CHECK-NEXT: .forward_progress: true +; GFX11-NEXT: .ieee_mode: false +; CHECK-NEXT: .image_op: false +; CHECK-NEXT: .lds_size: 0 +; CHECK-NEXT: .mem_ordered: true +; CHECK-NEXT: .scratch_en: false +; CHECK-NEXT: .scratch_memory_size: 0 +; CHECK-NEXT: .sgpr_count: 0x22 +; CHECK-NEXT: .sgpr_limit: 0x6a +; CHECK-NEXT: .threadgroup_dimensions: +; CHECK-NEXT: - 0x1 +; CHECK-NEXT: - 0x400 +; CHECK-NEXT: - 0x1 +; CHECK-NEXT: .trap_present: false +; CHECK-NEXT: .user_data_reg_map: +; CHECK-NEXT: - 0x10000000 +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0 +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: - 0xffffffff +; CHECK-NEXT: .user_sgprs: 0x3 +; CHECK-NEXT: .vgpr_count: 0x2 +; CHECK-NEXT: .vgpr_limit: 0x100 +; CHECK-NEXT: .wavefront_size: 0x40 +; CHECK-NEXT: .wgp_mode: false +; CHECK-NEXT: .gs: +; CHECK-NEXT: .debug_mode: false +; CHECK-NOT: .entry_point: _amdgpu_gs_main +; CHECK-NEXT: .entry_point_symbol: gs_shader +; CHECK-NEXT: .forward_progress: true +; GFX11-NEXT: .ieee_mode: false +; CHECK-NEXT: .lds_size: 0x200 +; CHECK-NEXT: .mem_ordered: true +; CHECK-NEXT: .scratch_en: false +; CHECK-NEXT: .scratch_memory_size: 0 +; CHECK-NEXT: .sgpr_count: 0x1 +; CHECK-NEXT: .vgpr_count: 0x1 +; CHECK-NEXT: .wgp_mode: true +; CHECK-NEXT: .hs: +; CHECK-NEXT: .debug_mode: false +; CHECK-NOT: .entry_point: _amdgpu_hs_main +; CHECK-NEXT: .entry_point_symbol: hs_shader +; CHECK-NEXT: .forward_progress: true +; GFX11-NEXT: .ieee_mode: false +; CHECK-NEXT: .lds_size: 0x1000 +; CHECK-NEXT: .mem_ordered: true +; CHECK-NEXT: .scratch_en: false +; CHECK-NEXT: .scratch_memory_size: 0 +; CHECK-NEXT: .sgpr_count: 0x1 +; CHECK-NEXT: .vgpr_count: 0x1 +; CHECK-NEXT: .wgp_mode: true +; CHECK-NEXT: .ps: +; CHECK-NEXT: .debug_mode: false +; CHECK-NOT: .entry_point: _amdgpu_ps_main +; CHECK-NEXT: .entry_point_symbol: ps_shader +; CHECK-NEXT: .forward_progress: true +; GFX11-NEXT: .ieee_mode: false +; CHECK-NEXT: .lds_size: 0 +; CHECK-NEXT: .mem_ordered: true +; CHECK-NEXT: .scratch_en: false +; CHECK-NEXT: .scratch_memory_size: 0 +; CHECK-NEXT: .sgpr_count: 0x1 +; CHECK-NEXT: .vgpr_count: 0x1 +; CHECK-NEXT: .wgp_mode: true +; CHECK: .registers: {} +; CHECK:amdpal.version: +; CHECK-NEXT: - 0x3 +; CHECK-NEXT: - 0x6 +; CHECK-NEXT:... +; CHECK-NEXT: .end_amdgpu_pal_metadata + +define dllexport amdgpu_cs void @_amdgpu_cs_main(i32 inreg %arg1, i32 %arg2) #0 !lgc.shaderstage !1 { +.entry: + %i = call i64 @llvm.amdgcn.s.getpc() + %i1 = and i64 %i, -4294967296 + %i2 = zext i32 %arg1 to i64 + %i3 = or i64 %i1, %i2 + %i4 = inttoptr i64 %i3 to ptr addrspace(4) + %i5 = and i32 %arg2, 1023 + %i6 = lshr i32 %arg2, 10 + %i7 = and i32 %i6, 1023 + %i8 = add nuw nsw i32 %i7, %i5 + %i9 = load <4 x i32>, ptr addrspace(4) %i4, align 16 + %.idx = shl nuw nsw i32 %i8, 2 + call void @llvm.amdgcn.raw.buffer.store.i32(i32 1, <4 x i32> %i9, i32 %.idx, i32 0, i32 0) + ret void +} + +define dllexport amdgpu_ps void @ps_shader() #1 { + ret void +} + +@LDS.GS = external addrspace(3) global [1 x i32], align 4 + +define dllexport amdgpu_gs void @gs_shader() { + %ptr = getelementptr i32, ptr addrspace(3) @LDS.GS, i32 0 + store i32 0, ptr addrspace(3) %ptr, align 4 + ret void +} + +@LDS.HS = external addrspace(3) global [1024 x i32], align 4 + +define dllexport amdgpu_hs void @hs_shader() { + %ptr = getelementptr i32, ptr addrspace(3) @LDS.HS, i32 0 + store i32 0, ptr addrspace(3) %ptr, align 4 + ret void +} + +!amdgpu.pal.metadata.msgpack = !{!0} + +attributes #0 = { nounwind memory(readwrite) "target-features"=",+wavefrontsize64,+cumode" "amdgpu-dynamic-vgpr-block-size"="16" } + +attributes #1 = { nounwind memory(readwrite) "InitialPSInputAddr"="36983" "amdgpu-dynamic-vgpr-block-size"="16" } + +!0 = !{!"\82\B0amdpal.pipelines\91\8A\A4.api\A6Vulkan\B2.compute_registers\85\AB.tg_size_en\C3\AA.tgid_x_en\C2\AA.tgid_y_en\C2\AA.tgid_z_en\C2\AF.tidig_comp_cnt\01\B0.hardware_stages\81\A3.cs\8C\AF.checksum_value\CE\94D\D7\D0\AB.debug_mode\00\AB.float_mode\CC\C0\A9.image_op\C2\AC.mem_ordered\C3\AB.sgpr_limitj\B7.threadgroup_dimensions\93\01\CD\04\00\01\AD.trap_present\00\B2.user_data_reg_map\DC\00 \CE\10\00\00\00\CE\FF\FF\FF\FF\00\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\CE\FF\FF\FF\FF\AB.user_sgprs\03\AB.vgpr_limit\CD\01\00\AF.wavefront_size@\B7.internal_pipeline_hash\92\CF\E7\10k\A6:\A6%\F7\CF\B2\1F\1A\D4{\DA\E1T\AA.registers\80\A8.shaders\81\A8.compute\82\B0.api_shader_hash\92\CF\E9Zn7}\1E\B9\E7\00\B1.hardware_mapping\91\A3.cs\B0.spill_threshold\CE\FF\FF\FF\FF\A5.type\A2Cs\B0.user_data_limit\01\AF.xgl_cache_info\82\B3.128_bit_cache_hash\92\CF\B4X\B8\11[\A4\88P\CF\A0;\B0\AF\FF\B4\BE\C0\AD.llpc_version\A461.1\AEamdpal.version\92\03\06"} +!1 = !{i32 7} diff --git a/llvm/test/CodeGen/AMDGPU/pal-metadata-3.6.ll b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.6.ll index 830872a..d2f26e8 100644 --- a/llvm/test/CodeGen/AMDGPU/pal-metadata-3.6.ll +++ b/llvm/test/CodeGen/AMDGPU/pal-metadata-3.6.ll @@ -1,17 +1,14 @@ -; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1100 <%s | FileCheck %s --check-prefixes=CHECK,GFX11,NODVGPR -; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 <%s | FileCheck %s --check-prefixes=CHECK,NODVGPR -; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 -mattr=+dynamic-vgpr <%s | FileCheck %s --check-prefixes=CHECK,DVGPR +; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1100 <%s | FileCheck %s --check-prefixes=CHECK,GFX11 +; RUN: llc -mtriple=amdgcn--amdpal -mcpu=gfx1200 <%s | FileCheck %s --check-prefixes=CHECK ; CHECK-LABEL: {{^}}_amdgpu_cs_main: -; NODVGPR: ; TotalNumSgprs: 4 -; DVGPR: ; TotalNumSgprs: 34 +; CHECK: ; TotalNumSgprs: 4 ; CHECK: ; NumVgprs: 2 ; CHECK: .amdgpu_pal_metadata ; CHECK-NEXT: --- ; CHECK-NEXT: amdpal.pipelines: ; CHECK-NEXT: - .api: Vulkan ; CHECK-NEXT: .compute_registers: -; DVGPR-NEXT: .dynamic_vgpr_en: true ; CHECK-NEXT: .tg_size_en: true ; CHECK-NEXT: .tgid_x_en: false ; CHECK-NEXT: .tgid_y_en: false @@ -57,7 +54,6 @@ ; CHECK-NEXT: .cs: ; CHECK-NEXT: .checksum_value: 0x9444d7d0 ; CHECK-NEXT: .debug_mode: false -; DVGPR-NEXT: .dynamic_vgpr_saved_count: 0x70 ; CHECK-NOT: .entry_point: _amdgpu_cs_main ; CHECK-NEXT: .entry_point_symbol: _amdgpu_cs_main ; CHECK-NEXT: .excp_en: 0 @@ -69,8 +65,7 @@ ; CHECK-NEXT: .mem_ordered: true ; CHECK-NEXT: .scratch_en: false ; CHECK-NEXT: .scratch_memory_size: 0 -; NODVGPR-NEXT: .sgpr_count: 0x4 -; DVGPR-NEXT: .sgpr_count: 0x22 +; CHECK-NEXT: .sgpr_count: 0x4 ; CHECK-NEXT: .sgpr_limit: 0x6a ; CHECK-NEXT: .threadgroup_dimensions: ; CHECK-NEXT: - 0x1 diff --git a/llvm/test/CodeGen/NVPTX/convert-sm103a.ll b/llvm/test/CodeGen/NVPTX/convert-sm103a.ll new file mode 100644 index 0000000..54b4dd8 --- /dev/null +++ b/llvm/test/CodeGen/NVPTX/convert-sm103a.ll @@ -0,0 +1,297 @@ +; NOTE: Assertions have been autogenerated by utils/update_llc_test_checks.py UTC_ARGS: --version 6 +; RUN: llc < %s -mtriple=nvptx64 -mcpu=sm_100a -mattr=+ptx87 | FileCheck %s +; RUN: llc < %s -mtriple=nvptx64 -mcpu=sm_103a -mattr=+ptx87 | FileCheck %s +; RUN: %if ptxas-sm_100a && ptxas-isa-8.7 %{ llc < %s -mtriple=nvptx64 -mcpu=sm_100a -mattr=+ptx87 | %ptxas-verify -arch=sm_100a %} +; RUN: %if ptxas-sm_103a && ptxas-isa-8.7 %{ llc < %s -mtriple=nvptx64 -mcpu=sm_103a -mattr=+ptx87 | %ptxas-verify -arch=sm_103a %} + +; F16X2 conversions + +define <2 x half> @cvt_rs_f16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_f16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_f16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_f16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_f16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.f16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x half> @llvm.nvvm.ff2f16x2.rs(float %f1, float %f2, i32 %rbits) + ret <2 x half> %val +} + +define <2 x half> @cvt_rs_relu_f16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_f16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_relu_f16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_relu_f16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_relu_f16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.relu.f16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x half> @llvm.nvvm.ff2f16x2.rs.relu(float %f1, float %f2, i32 %rbits) + ret <2 x half> %val +} + +define <2 x half> @cvt_rs_sf_f16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_sf_f16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_sf_f16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_sf_f16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_sf_f16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.satfinite.f16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x half> @llvm.nvvm.ff2f16x2.rs.satfinite(float %f1, float %f2, i32 %rbits) + ret <2 x half> %val +} + +define <2 x half> @cvt_rs_relu_sf_f16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_sf_f16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_relu_sf_f16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_relu_sf_f16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_relu_sf_f16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.relu.satfinite.f16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x half> @llvm.nvvm.ff2f16x2.rs.relu.satfinite(float %f1, float %f2, i32 %rbits) + ret <2 x half> %val +} + +; BF16X2 conversions + +define <2 x bfloat> @cvt_rs_bf16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_bf16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_bf16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_bf16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_bf16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.bf16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x bfloat> @llvm.nvvm.ff2bf16x2.rs(float %f1, float %f2, i32 %rbits) + ret <2 x bfloat> %val +} + +define <2 x bfloat> @cvt_rs_relu_bf16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_bf16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_relu_bf16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_relu_bf16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_relu_bf16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.relu.bf16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x bfloat> @llvm.nvvm.ff2bf16x2.rs.relu(float %f1, float %f2, i32 %rbits) + ret <2 x bfloat> %val +} + +define <2 x bfloat> @cvt_rs_sf_bf16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_sf_bf16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_sf_bf16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_sf_bf16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_sf_bf16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.satfinite.bf16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x bfloat> @llvm.nvvm.ff2bf16x2.rs.satfinite(float %f1, float %f2, i32 %rbits) + ret <2 x bfloat> %val +} + +define <2 x bfloat> @cvt_rs_relu_sf_bf16x2_f32(float %f1, float %f2, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_sf_bf16x2_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<5>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.b32 %r1, [cvt_rs_relu_sf_bf16x2_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r2, [cvt_rs_relu_sf_bf16x2_f32_param_1]; +; CHECK-NEXT: ld.param.b32 %r3, [cvt_rs_relu_sf_bf16x2_f32_param_2]; +; CHECK-NEXT: cvt.rs.relu.satfinite.bf16x2.f32 %r4, %r1, %r2, %r3; +; CHECK-NEXT: st.param.b32 [func_retval0], %r4; +; CHECK-NEXT: ret; + %val = call <2 x bfloat> @llvm.nvvm.ff2bf16x2.rs.relu.satfinite(float %f1, float %f2, i32 %rbits) + ret <2 x bfloat> %val +} + +; F8X4 conversions + +define <4 x i8> @cvt_rs_sf_e4m3x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_sf_e4m3x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_sf_e4m3x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_sf_e4m3x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.satfinite.e4m3x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e4m3x4.rs.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +define <4 x i8> @cvt_rs_relu_sf_e4m3x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_sf_e4m3x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_relu_sf_e4m3x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_relu_sf_e4m3x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.relu.satfinite.e4m3x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e4m3x4.rs.relu.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +define <4 x i8> @cvt_rs_sf_e5m2x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_sf_e5m2x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_sf_e5m2x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_sf_e5m2x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.satfinite.e5m2x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e5m2x4.rs.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +define <4 x i8> @cvt_rs_relu_sf_e5m2x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_sf_e5m2x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_relu_sf_e5m2x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_relu_sf_e5m2x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.relu.satfinite.e5m2x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e5m2x4.rs.relu.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +; F6X4 conversions + +define <4 x i8> @cvt_rs_sf_e2m3x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_sf_e2m3x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_sf_e2m3x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_sf_e2m3x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.satfinite.e2m3x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e2m3x4.rs.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +define <4 x i8> @cvt_rs_relu_sf_e2m3x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_sf_e2m3x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_relu_sf_e2m3x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_relu_sf_e2m3x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.relu.satfinite.e2m3x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e2m3x4.rs.relu.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +define <4 x i8> @cvt_rs_sf_e3m2x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_sf_e3m2x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_sf_e3m2x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_sf_e3m2x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.satfinite.e3m2x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e3m2x4.rs.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +define <4 x i8> @cvt_rs_relu_sf_e3m2x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_sf_e3m2x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_relu_sf_e3m2x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_relu_sf_e3m2x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.relu.satfinite.e3m2x4.f32 %r6, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call <4 x i8> @llvm.nvvm.f32x4.to.e3m2x4.rs.relu.satfinite(<4 x float> %fvec, i32 %rbits) + ret <4 x i8> %val +} + +; F4X4 conversions + +define i16 @cvt_rs_sf_e2m1x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_sf_e2m1x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b16 %rs<2>; +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_sf_e2m1x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_sf_e2m1x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.satfinite.e2m1x4.f32 %rs1, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: cvt.u32.u16 %r6, %rs1; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call i16 @llvm.nvvm.f32x4.to.e2m1x4.rs.satfinite(<4 x float> %fvec, i32 %rbits) + ret i16 %val +} + +define i16 @cvt_rs_relu_sf_e2m1x4_f32(<4 x float> %fvec, i32 %rbits) { +; CHECK-LABEL: cvt_rs_relu_sf_e2m1x4_f32( +; CHECK: { +; CHECK-NEXT: .reg .b16 %rs<2>; +; CHECK-NEXT: .reg .b32 %r<7>; +; CHECK-EMPTY: +; CHECK-NEXT: // %bb.0: +; CHECK-NEXT: ld.param.v4.b32 {%r1, %r2, %r3, %r4}, [cvt_rs_relu_sf_e2m1x4_f32_param_0]; +; CHECK-NEXT: ld.param.b32 %r5, [cvt_rs_relu_sf_e2m1x4_f32_param_1]; +; CHECK-NEXT: cvt.rs.relu.satfinite.e2m1x4.f32 %rs1, {%r1, %r2, %r3, %r4}, %r5; +; CHECK-NEXT: cvt.u32.u16 %r6, %rs1; +; CHECK-NEXT: st.param.b32 [func_retval0], %r6; +; CHECK-NEXT: ret; + %val = call i16 @llvm.nvvm.f32x4.to.e2m1x4.rs.relu.satfinite(<4 x float> %fvec, i32 %rbits) + ret i16 %val +} diff --git a/llvm/test/CodeGen/NVPTX/wmma-ptx87-sm120a.py b/llvm/test/CodeGen/NVPTX/wmma-ptx87-sm120a.py index ae781df..40055ae 100644 --- a/llvm/test/CodeGen/NVPTX/wmma-ptx87-sm120a.py +++ b/llvm/test/CodeGen/NVPTX/wmma-ptx87-sm120a.py @@ -2,7 +2,7 @@ # RUN: %python %s --ptx=87 --gpu-arch=120 --aa > %t-ptx87-sm_120a.ll # RUN: llc < %t-ptx87-sm_120a.ll -mtriple=nvptx64 -mcpu=sm_120a -mattr=+ptx87 \ # RUN: | FileCheck %t-ptx87-sm_120a.ll -# RUN: %if ptxas-12.7 %{ \ +# RUN: %if ptxas-sm_120a && ptxas-isa-8.7 %{ \ # RUN: llc < %t-ptx87-sm_120a.ll -mtriple=nvptx64 -mcpu=sm_120a -mattr=+ptx87 \ # RUN: | %ptxas-verify -arch=sm_120a \ # RUN: %} diff --git a/llvm/test/CodeGen/NVPTX/wmma.py b/llvm/test/CodeGen/NVPTX/wmma.py index 6d73bce..8427ae4 100644 --- a/llvm/test/CodeGen/NVPTX/wmma.py +++ b/llvm/test/CodeGen/NVPTX/wmma.py @@ -90,6 +90,21 @@ class MMAFrag: "m16n8k32:b:s8": 2, "m16n8k32:c:s32": 4, "m16n8k32:d:s32": 4, + # e4m3/e5m2/e3m2/e2m3/e2m1 -> f16/f32 @ m16n8k16/m16n8k32 + "m16n8k16:a:e4m3": 2, + "m16n8k16:a:e5m2": 2, + "m16n8k32:a:e4m3": 4, + "m16n8k32:a:e5m2": 4, + "m16n8k32:a:e3m2": 4, + "m16n8k32:a:e2m3": 4, + "m16n8k32:a:e2m1": 4, + "m16n8k16:b:e4m3": 1, + "m16n8k16:b:e5m2": 1, + "m16n8k32:b:e4m3": 2, + "m16n8k32:b:e5m2": 2, + "m16n8k32:b:e3m2": 2, + "m16n8k32:b:e2m3": 2, + "m16n8k32:b:e2m1": 2, # mma sp "m16n8k32:a:bf16": 4, "m16n8k32:a:f16": 4, @@ -182,6 +197,18 @@ class MMAFrag: "m8n8k4:b:f64": 1, "m8n8k4:c:f64": 2, "m8n8k4:d:f64": 2, + "m16n8k4:a:f64": 2, + "m16n8k4:b:f64": 1, + "m16n8k4:c:f64": 4, + "m16n8k4:d:f64": 4, + "m16n8k8:a:f64": 4, + "m16n8k8:b:f64": 2, + "m16n8k8:c:f64": 4, + "m16n8k8:d:f64": 4, + "m16n8k16:a:f64": 8, + "m16n8k16:b:f64": 4, + "m16n8k16:c:f64": 4, + "m16n8k16:d:f64": 4, # tf32 -> s32 @ m16n16k8 "m16n16k8:a:tf32": 4, "m16n16k8:b:tf32": 4, @@ -324,7 +351,9 @@ def get_wmma_ops(): def get_mma_ops(): return ( - make_mma_ops(["m8n8k4"], ["f64"], [], ["f64"], []) + make_mma_ops( + ["m8n8k4", "m16n8k4", "m16n8k8", "m16n8k16"], ["f64"], [], ["f64"], [] + ) + make_mma_ops(["m16n8k4", "m16n8k8"], ["tf32"], [], ["f32"], []) + make_mma_ops(["m16n8k16", "m16n8k8"], ["bf16"], [], ["f32"], []) + make_mma_ops( @@ -341,6 +370,20 @@ def get_mma_ops(): ["m8n8k32", "m16n8k32", "m16n8k64"], ["s4", "u4"], ["s4", "u4"], ["s32"], [] ) + make_mma_ops(["m8n8k128", "m16n8k128", "m16n8k256"], ["b1"], [], ["s32"], []) + + make_mma_ops( + ["m16n8k16"], + ["e4m3", "e5m2"], + ["e4m3", "e5m2"], + ["f16", "f32"], + ["f16", "f32"], + ) + + make_mma_ops( + ["m16n8k32"], + ["e4m3", "e5m2", "e3m2", "e2m3", "e2m1"], + ["e4m3", "e5m2", "e3m2", "e2m3", "e2m1"], + ["f16", "f32"], + ["f16", "f32"], + ) ) @@ -492,7 +535,7 @@ def is_wmma_variant_supported(op, layout_a, layout_b, rnd, satf): return True -def is_mma_variant_supported(op, layout_a, layout_b, satf): +def is_mma_variant_supported(op, layout_a, layout_b, kind, satf): if not ( is_type_supported(op.a.mma_type.ptx_type) and is_mma_geom_supported(op.a.geom) ): @@ -516,13 +559,53 @@ def is_mma_variant_supported(op, layout_a, layout_b, satf): ): return False + if ( + op.a.geom != "m8n8k4" + and op.a.mma_type.ptx_type == "f64" + and (ptx_version < 78 or gpu_arch < 90) + ): + return False + # C and D type must be the same - if op.a.geom == "m16n8k16" and op.c.mma_type.ptx_type != op.d.mma_type.ptx_type: + if ( + op.a.geom in ["m16n8k16", "m16n8k32"] + and op.c.mma_type.ptx_type != op.d.mma_type.ptx_type + ): + return False + + if ( + op.a.geom in ["m16n8k16", "m16n8k32"] + and any( + x in ["e4m3", "e5m2"] + for x in (op.a.mma_type.ptx_type, op.b.mma_type.ptx_type) + ) + and ptx_version < 87 + ): + return False + + if kind != "" and not (ptx_version >= 87 and gpu_arch >= 120 and aa): + return False + + if kind != "" and ( + op.a.geom != "m16n8k32" + or op.a.mma_type.ptx_type not in ["e4m3", "e5m2", "e3m2", "e2m3", "e2m1"] + ): + return False + + if ( + kind == "" + and op.a.geom in ["m16n8k16", "m16n8k32"] + and any( + x in ["e3m2", "e2m3", "e2m1"] + for x in (op.a.mma_type.ptx_type, op.b.mma_type.ptx_type) + ) + ): return False # Require row/col layout for all MMA except m8n8k4 on FP16 if not (op.a.geom == "m8n8k4" and op.a.mma_type.ptx_type == "f16"): return layout_a == "row" and layout_b == "col" + return True @@ -937,7 +1020,12 @@ define ${ret_ty} @test_${function}( """ test_params = params - test_params["intrinsic"] = Template(intrinsic_template).substitute(params) + test_params["intrinsic"] = ( + Template(intrinsic_template) + .substitute(params) + .replace("::", ".") + .replace("_", ".") + ) test_params["function"] = test_params["intrinsic"].replace(".", "_") test_params["instruction"] = Template(instruction_template).substitute(params) test_params["ret_ty"] = make_wmma_ld_ret_ty(op.d) @@ -1002,16 +1090,20 @@ def gen_wmma_mma_tests(): def gen_mma_tests(): - mma_intrinsic_template = "llvm.nvvm.mma${b1op}.${geom}.${alayout}.${blayout}${satf}.${intrinsic_signature}" - mma_instruction_template = "mma.sync${aligned}.${geom}.${alayout}.${blayout}${satf}.${ptx_signature}${b1op}" + mma_intrinsic_template = "llvm.nvvm.mma${b1op}.${geom}.${alayout}.${blayout}${kind}${satf}.${intrinsic_signature}" + mma_instruction_template = "mma.sync${aligned}.${geom}.${alayout}.${blayout}${kind}${satf}.${ptx_signature}${b1op}" generated_items = [] - for op, alayout, blayout, satf in product( - get_mma_ops(), ["row", "col"], ["row", "col"], [".satfinite", ""] + for op, alayout, blayout, kind, satf in product( + get_mma_ops(), + ["row", "col"], + ["row", "col"], + ["", ".kind::f8f6f4"], + [".satfinite", ""], ): - if not is_mma_variant_supported(op, alayout, blayout, satf): + if not is_mma_variant_supported(op, alayout, blayout, kind, satf): continue for b1op in get_b1_ops(op.a.mma_type.ptx_type): @@ -1024,6 +1116,7 @@ def gen_mma_tests(): "satf": satf, "geom": op.a.geom, "b1op": b1op, + "kind": kind, } intrinsic_template = mma_intrinsic_template @@ -1105,9 +1198,9 @@ def is_mma_sp_variant_supported(op, metadata, kind, satf): ): return False - # C and D type must be the same for m16n8k16/m16n8k32 + # C and D type must be the same for m16n8k16/m16n8k32/m16n8k64 if ( - op.a.geom in ["m16n8k16", "m16n8k32"] + op.a.geom in ["m16n8k16", "m16n8k32", "m16n8k64"] and op.c.mma_type.ptx_type != op.d.mma_type.ptx_type ): return False diff --git a/llvm/test/CodeGen/SPIRV/hlsl-resources/test_counters.ll b/llvm/test/CodeGen/SPIRV/hlsl-resources/test_counters.ll new file mode 100644 index 0000000..b178a56 --- /dev/null +++ b/llvm/test/CodeGen/SPIRV/hlsl-resources/test_counters.ll @@ -0,0 +1,65 @@ +; RUN: llc -O0 -verify-machineinstrs -mtriple=spirv-vulkan-library %s -o - | FileCheck %s +; RUN: %if spirv-tools %{ llc -O0 -mtriple=spirv-vulkan-library %s -o - -filetype=obj | spirv-val --target-env vulkan1.3 %} + +; ModuleID = 'test_counters.hlsl' +source_filename = "test_counters.hlsl" + +; CHECK: OpCapability Int8 +; CHECK-DAG: OpName [[OutputBuffer:%[0-9]+]] "OutputBuffer" +; CHECK-DAG: OpName [[InputBuffer:%[0-9]+]] "InputBuffer" +; CHECK-DAG: OpName [[OutputBufferCounter:%[0-9]+]] "OutputBuffer.counter" +; CHECK-DAG: OpName [[InputBufferCounter:%[0-9]+]] "InputBuffer.counter" +; CHECK-DAG: OpDecorate [[OutputBuffer]] DescriptorSet 0 +; CHECK-DAG: OpDecorate [[OutputBuffer]] Binding 10 +; CHECK-DAG: OpDecorate [[OutputBufferCounter]] DescriptorSet 0 +; CHECK-DAG: OpDecorate [[OutputBufferCounter]] Binding 0 +; CHECK-DAG: OpDecorate [[InputBuffer]] DescriptorSet 0 +; CHECK-DAG: OpDecorate [[InputBuffer]] Binding 1 +; CHECK-DAG: OpDecorate [[InputBufferCounter]] DescriptorSet 0 +; CHECK-DAG: OpDecorate [[InputBufferCounter]] Binding 2 +; CHECK-DAG: [[int:%[0-9]+]] = OpTypeInt 32 0 +; CHECK-DAG: [[zero:%[0-9]+]] = OpConstant [[int]] 0{{$}} +; CHECK-DAG: [[one:%[0-9]+]] = OpConstant [[int]] 1{{$}} +; CHECK-DAG: [[minus_one:%[0-9]+]] = OpConstant [[int]] 4294967295 +; CHECK: [[OutputBufferHandle:%[0-9]+]] = OpCopyObject {{%[0-9]+}} [[OutputBuffer]] +; CHECK: [[InputBufferHandle:%[0-9]+]] = OpCopyObject {{%[0-9]+}} [[InputBuffer]] +; CHECK: [[InputCounterAC:%[0-9]+]] = OpAccessChain {{%[0-9]+}} [[InputBufferCounter]] [[zero]] +; CHECK: [[dec:%[0-9]+]] = OpAtomicIAdd [[int]] [[InputCounterAC]] [[one]] [[zero]] [[minus_one]] +; CHECK: [[iadd:%[0-9]+]] = OpIAdd [[int]] [[dec]] [[minus_one]] +; CHECK: [[OutputCounterAC:%[0-9]+]] = OpAccessChain {{%[0-9]+}} [[OutputBufferCounter]] [[zero]] +; CHECK: [[inc:%[0-9]+]] = OpAtomicIAdd [[int]] [[OutputCounterAC]] [[one]] [[zero]] [[one]] +; CHECK: [[InputAC:%[0-9]+]] = OpAccessChain {{%[0-9]+}} [[InputBufferHandle]] [[zero]] [[iadd]] +; CHECK: [[load:%[0-9]+]] = OpLoad {{%[0-9]+}} [[InputAC]] +; CHECK: [[OutputAC:%[0-9]+]] = OpAccessChain {{%[0-9]+}} [[OutputBufferHandle]] [[zero]] [[inc]] +; CHECK: OpStore [[OutputAC]] [[load]] + + +target triple = "spirv1.6-unknown-vulkan1.3-compute" + +@.str = private unnamed_addr constant [13 x i8] c"OutputBuffer\00" +@.str.2 = private unnamed_addr constant [12 x i8] c"InputBuffer\00" + +define void @main() #0 { +entry: + %0 = call target("spirv.VulkanBuffer", [0 x float], 12, 1) @llvm.spv.resource.handlefrombinding.tspirv.VulkanBuffer_a0f32_12_1t(i32 0, i32 10, i32 1, i32 0, ptr @.str) + %1 = call target("spirv.VulkanBuffer", i32, 12, 1) @llvm.spv.resource.counterhandlefromimplicitbinding.tspirv.VulkanBuffer_i32_12_1t.tspirv.VulkanBuffer_a0f32_12_1t(target("spirv.VulkanBuffer", [0 x float], 12, 1) %0, i32 0, i32 0) + %2 = call target("spirv.VulkanBuffer", [0 x float], 12, 1) @llvm.spv.resource.handlefromimplicitbinding.tspirv.VulkanBuffer_a0f32_12_1t(i32 1, i32 0, i32 1, i32 0, ptr @.str.2) + %3 = call target("spirv.VulkanBuffer", i32, 12, 1) @llvm.spv.resource.counterhandlefromimplicitbinding.tspirv.VulkanBuffer_i32_12_1t.tspirv.VulkanBuffer_a0f32_12_1t(target("spirv.VulkanBuffer", [0 x float], 12, 1) %2, i32 2, i32 0) + %4 = call i32 @llvm.spv.resource.updatecounter.tspirv.VulkanBuffer_i32_12_1t(target("spirv.VulkanBuffer", i32, 12, 1) %3, i8 -1) + %5 = call i32 @llvm.spv.resource.updatecounter.tspirv.VulkanBuffer_i32_12_1t(target("spirv.VulkanBuffer", i32, 12, 1) %1, i8 1) + %6 = call ptr addrspace(11) @llvm.spv.resource.getpointer.p11.tspirv.VulkanBuffer_a0f32_12_1t(target("spirv.VulkanBuffer", [0 x float], 12, 1) %2, i32 %4) + %7 = load float, ptr addrspace(11) %6 + %8 = call ptr addrspace(11) @llvm.spv.resource.getpointer.p11.tspirv.VulkanBuffer_a0f32_12_1t(target("spirv.VulkanBuffer", [0 x float], 12, 1) %0, i32 %5) + store float %7, ptr addrspace(11) %8 + ret void +} + +declare target("spirv.VulkanBuffer", [0 x float], 12, 1) @llvm.spv.resource.handlefrombinding.tspirv.VulkanBuffer_a0f32_12_1t(i32, i32, i32, i32, ptr) #1 +declare target("spirv.VulkanBuffer", i32, 12, 1) @llvm.spv.resource.counterhandlefromimplicitbinding.tspirv.VulkanBuffer_i32_12_1t.tspirv.VulkanBuffer_a0f32_12_1t(target("spirv.VulkanBuffer", [0 x float], 12, 1), i32, i32) #1 +declare target("spirv.VulkanBuffer", [0 x float], 12, 1) @llvm.spv.resource.handlefromimplicitbinding.tspirv.VulkanBuffer_a0f32_12_1t(i32, i32, i32, i32, ptr) #1 +declare i32 @llvm.spv.resource.updatecounter.tspirv.VulkanBuffer_i32_12_1t(target("spirv.VulkanBuffer", i32, 12, 1), i8) #2 +declare ptr addrspace(11) @llvm.spv.resource.getpointer.p11.tspirv.VulkanBuffer_a0f32_12_1t(target("spirv.VulkanBuffer", [0 x float], 12, 1), i32) #1 + +attributes #0 = { "hlsl.shader"="compute" "hlsl.numthreads"="1,1,1" } +attributes #1 = { memory(none) } +attributes #2 = { memory(argmem: readwrite, inaccessiblemem: readwrite) } diff --git a/llvm/test/CodeGen/X86/GlobalISel/legalize-phi.mir b/llvm/test/CodeGen/X86/GlobalISel/legalize-phi.mir index 31de686..92e4588 100644 --- a/llvm/test/CodeGen/X86/GlobalISel/legalize-phi.mir +++ b/llvm/test/CodeGen/X86/GlobalISel/legalize-phi.mir @@ -148,21 +148,21 @@ body: | ; CHECK-NEXT: {{ $}} ; CHECK-NEXT: [[COPY:%[0-9]+]]:_(s32) = COPY $edi ; CHECK-NEXT: [[COPY1:%[0-9]+]]:_(s32) = COPY $esi + ; CHECK-NEXT: [[TRUNC:%[0-9]+]]:_(s1) = G_TRUNC [[COPY1]](s32) ; CHECK-NEXT: [[COPY2:%[0-9]+]]:_(s32) = COPY $edx + ; CHECK-NEXT: [[TRUNC2:%[0-9]+]]:_(s1) = G_TRUNC [[COPY2]](s32) ; CHECK-NEXT: [[C:%[0-9]+]]:_(s32) = G_CONSTANT i32 0 ; CHECK-NEXT: [[ICMP:%[0-9]+]]:_(s8) = G_ICMP intpred(sgt), [[COPY]](s32), [[C]] - ; CHECK-NEXT: [[TRUNC:%[0-9]+]]:_(s1) = G_TRUNC [[ICMP]](s8) - ; CHECK-NEXT: [[TRUNC1:%[0-9]+]]:_(s8) = G_TRUNC [[COPY1]](s32) - ; CHECK-NEXT: G_BRCOND [[TRUNC]](s1), %bb.2 + ; CHECK-NEXT: [[TRUNC1:%[0-9]+]]:_(s1) = G_TRUNC [[ICMP]](s8) + ; CHECK-NEXT: G_BRCOND [[TRUNC1]](s1), %bb.2 ; CHECK-NEXT: {{ $}} ; CHECK-NEXT: bb.1.cond.false: ; CHECK-NEXT: successors: %bb.2(0x80000000) ; CHECK-NEXT: {{ $}} - ; CHECK-NEXT: [[TRUNC2:%[0-9]+]]:_(s8) = G_TRUNC [[COPY2]](s32) - ; CHECK-NEXT: {{ $}} ; CHECK-NEXT: bb.2.cond.end: - ; CHECK-NEXT: [[PHI:%[0-9]+]]:_(s8) = G_PHI [[TRUNC2]](s8), %bb.1, [[TRUNC1]](s8), %bb.0 - ; CHECK-NEXT: $al = COPY [[PHI]](s8) + ; CHECK-NEXT: [[PHI:%[0-9]+]]:_(s1) = G_PHI [[TRUNC2]](s1), %bb.1, [[TRUNC]](s1), %bb.0 + ; CHECK-NEXT: [[EXT:%[0-9]+]]:_(s8) = G_ANYEXT [[PHI]](s1) + ; CHECK-NEXT: $al = COPY [[EXT]](s8) ; CHECK-NEXT: RET 0, implicit $al bb.1.entry: successors: %bb.3(0x40000000), %bb.2(0x40000000) diff --git a/llvm/test/CodeGen/X86/GlobalISel/legalize-undef-vec-scaling.mir b/llvm/test/CodeGen/X86/GlobalISel/legalize-undef-vec-scaling.mir new file mode 100644 index 0000000..b02832b --- /dev/null +++ b/llvm/test/CodeGen/X86/GlobalISel/legalize-undef-vec-scaling.mir @@ -0,0 +1,32 @@ +# RUN: llc -mtriple=x86_64-linux-gnu -mattr=avx2 -run-pass=legalizer -global-isel-abort=2 -pass-remarks-missed='gisel*' %s -o - | FileCheck %s --check-prefixes=CHECK,AVX2 +# RUN: llc -mtriple=x86_64-linux-gnu -mattr=sse2 -run-pass=legalizer -global-isel-abort=2 -pass-remarks-missed='gisel*' %s -o - | FileCheck %s --check-prefixes=CHECK,SSE2 +# RUN: llc -mtriple=x86_64-linux-gnu -mattr=avx512f -run-pass=legalizer -global-isel-abort=2 -pass-remarks-missed='gisel*' %s -o - | FileCheck %s --check-prefixes=CHECK,AVX512F + + +--- +name: test_basic_g_implicit_def_v8i64 +body: | + bb.0: + ; CHECK-LABEL: name: test_basic_g_implicit_def_v8i64 + ; AVX512F: {{%[0-9]+}}:_(<8 x s64>) = G_IMPLICIT_DEF + ; AVX2: [[DEF_AVX2:%[0-9]+]]:_(<4 x s64>) = G_IMPLICIT_DEF + ; AVX2-NEXT: {{%[0-9]+}}:_(<8 x s64>) = G_CONCAT_VECTORS [[DEF_AVX2]](<4 x s64>), [[DEF_AVX2]](<4 x s64>) + ; SSE2: [[DEF_SSE2:%[0-9]+]]:_(<2 x s64>) = G_IMPLICIT_DEF + ; SSE2-NEXT: {{%[0-9]+}}:_(<8 x s64>) = G_CONCAT_VECTORS [[DEF_SSE2]](<2 x s64>), [[DEF_SSE2]](<2 x s64>), [[DEF_SSE2]](<2 x s64>), [[DEF_SSE2]](<2 x s64>) + %0:_(<8 x s64>) = G_IMPLICIT_DEF + RET 0, implicit %0 +... + +--- +name: test_g_implicit_def_cample_size +body: | + bb.1: + ; CHECK-LABEL: name: test_g_implicit_def_cample_size + ; AVX512: {{%[0-9]+}}:_(<8 x s64>) = G_IMPLICIT_DEF + ; AVX2: {{%[0-9]+}}:_(<4 x s64>) = G_IMPLICIT_DEF + ; SSE2: {{%[0-9]+}}:_(<2 x s64>) = G_IMPLICIT_DEF + %0:_(<5 x s63>) = G_IMPLICIT_DEF + RET 0, implicit %0 +... + + diff --git a/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier-vec256.mir b/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier-vec256.mir new file mode 100644 index 0000000..254c1b6 --- /dev/null +++ b/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier-vec256.mir @@ -0,0 +1,23 @@ +# RUN: llc -mtriple=x86_64-linux-gnu -mattr=+avx -run-pass=instruction-select -verify-machineinstrs %s -o - | FileCheck %s + +--- +name: select_cfb_vec256 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: vecr, preferred-register: '', flags: [ ] } + - { id: 1, class: vecr, preferred-register: '', flags: [ ] } +body: | + bb.0: + liveins: $ymm0 + + ; CHECK-LABEL: name: select_cfb_vec256 + ; CHECK: [[COPY:%[0-9]+]]:vr256 = COPY $ymm0 + ; CHECK-NOT: G_CONSTANT_FOLD_BARRIER + ; CHECK-NEXT: $ymm1 = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $ymm1 + %0:vecr(<8 x s32>) = COPY $ymm0 + %1:vecr(<8 x s32>) = G_CONSTANT_FOLD_BARRIER %0 + $ymm1 = COPY %1(<8 x s32>) + RET 0, implicit $ymm1 +... diff --git a/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier-vec512.mir b/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier-vec512.mir new file mode 100644 index 0000000..3da354b --- /dev/null +++ b/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier-vec512.mir @@ -0,0 +1,23 @@ +# RUN: llc -mtriple=x86_64-linux-gnu -mattr=+avx512f -run-pass=instruction-select -verify-machineinstrs %s -o - | FileCheck %s + +--- +name: select_cfb_vec512 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: vecr, preferred-register: '', flags: [ ] } + - { id: 1, class: vecr, preferred-register: '', flags: [ ] } +body: | + bb.0: + liveins: $zmm0 + + ; CHECK-LABEL: name: select_cfb_vec512 + ; CHECK: [[COPY:%[0-9]+]]:vr512 = COPY $zmm0 + ; CHECK-NOT: G_CONSTANT_FOLD_BARRIER + ; CHECK-NEXT: $zmm1 = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $zmm1 + %0:vecr(<8 x s64>) = COPY $zmm0 + %1:vecr(<8 x s64>) = G_CONSTANT_FOLD_BARRIER %0 + $zmm1 = COPY %1(<8 x s64>) + RET 0, implicit $zmm1 +... diff --git a/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier.mir b/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier.mir new file mode 100644 index 0000000..fa012f9 --- /dev/null +++ b/llvm/test/CodeGen/X86/GlobalISel/select-constant-fold-barrier.mir @@ -0,0 +1,77 @@ +# RUN: llc -mtriple=x86_64-linux-gnu -run-pass=instruction-select -verify-machineinstrs %s -o - | FileCheck %s + + +--- +name: select_cfb_scalar_s32 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: gpr, preferred-register: '', flags: [ ] } + - { id: 1, class: gpr, preferred-register: '', flags: [ ] } +liveins: +fixedStack: +stack: +constants: +body: | + bb.0: + liveins: $edi + + ; CHECK-LABEL: name: select_cfb_scalar_s32 + ; CHECK: [[COPY:%[0-9]+]]:gr32 = COPY $edi + ; CHECK-NOT: G_CONSTANT_FOLD_BARRIER + ; CHECK-NEXT: $eax = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $eax + %0:gpr(s32) = COPY $edi + %1:gpr(s32) = G_CONSTANT_FOLD_BARRIER %0 + $eax = COPY %1(s32) + RET 0, implicit $eax +... + +--- +name: select_cfb_scalar_s64 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: gpr, preferred-register: '', flags: [ ] } + - { id: 1, class: gpr, preferred-register: '', flags: [ ] } +liveins: +fixedStack: +stack: +constants: +body: | + bb.0: + liveins: $rdi + + ; CHECK-LABEL: name: select_cfb_scalar_s64 + ; CHECK: [[COPY:%[0-9]+]]:gr64 = COPY $rdi + ; CHECK-NOT: G_CONSTANT_FOLD_BARRIER + ; CHECK-NEXT: $rax = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $rax + %0:gpr(s64) = COPY $rdi + %1:gpr(s64) = G_CONSTANT_FOLD_BARRIER %0 + $rax = COPY %1(s64) + RET 0, implicit $rax +... + + +--- +name: select_cfb_vec128 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: vecr, preferred-register: '', flags: [ ] } + - { id: 1, class: vecr, preferred-register: '', flags: [ ] } +body: | + bb.0: + liveins: $xmm0 + + ; CHECK-LABEL: name: select_cfb_vec128 + ; CHECK: [[COPY:%[0-9]+]]:vr128 = COPY $xmm0 + ; CHECK-NOT: G_CONSTANT_FOLD_BARRIER + ; CHECK-NEXT: $xmm1 = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $xmm1 + %0:vecr(<4 x s32>) = COPY $xmm0 + %1:vecr(<4 x s32>) = G_CONSTANT_FOLD_BARRIER %0 + $xmm1 = COPY %1(<4 x s32>) + RET 0, implicit $xmm1 +... diff --git a/llvm/test/CodeGen/X86/GlobalISel/select-freeze-vec256.mir b/llvm/test/CodeGen/X86/GlobalISel/select-freeze-vec256.mir new file mode 100644 index 0000000..11251e4 --- /dev/null +++ b/llvm/test/CodeGen/X86/GlobalISel/select-freeze-vec256.mir @@ -0,0 +1,23 @@ +# RUN: llc -mtriple=x86_64-linux-gnu -mattr=+avx -run-pass=instruction-select -verify-machineinstrs %s -o - | FileCheck %s + +--- +name: select_freeze_vec256 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: vecr, preferred-register: '', flags: [ ] } + - { id: 1, class: vecr, preferred-register: '', flags: [ ] } +body: | + bb.0: + liveins: $ymm0 + + ; CHECK-LABEL: name: select_freeze_vec256 + ; CHECK: [[COPY:%[0-9]+]]:vr256 = COPY $ymm0 + ; CHECK-NOT: G_FREEZE + ; CHECK-NEXT: $ymm1 = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $ymm1 + %0:vecr(<8 x s32>) = COPY $ymm0 + %1:vecr(<8 x s32>) = G_FREEZE %0 + $ymm1 = COPY %1(<8 x s32>) + RET 0, implicit $ymm1 +... diff --git a/llvm/test/CodeGen/X86/GlobalISel/select-freeze-vec512.mir b/llvm/test/CodeGen/X86/GlobalISel/select-freeze-vec512.mir new file mode 100644 index 0000000..bcf299a --- /dev/null +++ b/llvm/test/CodeGen/X86/GlobalISel/select-freeze-vec512.mir @@ -0,0 +1,23 @@ +# RUN: llc -mtriple=x86_64-linux-gnu -mattr=+avx512f -run-pass=instruction-select -verify-machineinstrs %s -o - | FileCheck %s + +--- +name: select_freeze_vec512 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: vecr, preferred-register: '', flags: [ ] } + - { id: 1, class: vecr, preferred-register: '', flags: [ ] } +body: | + bb.0: + liveins: $zmm0 + + ; CHECK-LABEL: name: select_freeze_vec512 + ; CHECK: [[COPY:%[0-9]+]]:vr512 = COPY $zmm0 + ; CHECK-NOT: G_FREEZE + ; CHECK-NEXT: $zmm1 = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $zmm1 + %0:vecr(<8 x s64>) = COPY $zmm0 + %1:vecr(<8 x s64>) = G_FREEZE %0 + $zmm1 = COPY %1(<8 x s64>) + RET 0, implicit $zmm1 +... diff --git a/llvm/test/CodeGen/X86/GlobalISel/select-freeze.mir b/llvm/test/CodeGen/X86/GlobalISel/select-freeze.mir new file mode 100644 index 0000000..cf5ad47 --- /dev/null +++ b/llvm/test/CodeGen/X86/GlobalISel/select-freeze.mir @@ -0,0 +1,77 @@ +# RUN: llc -mtriple=x86_64-linux-gnu -run-pass=instruction-select -verify-machineinstrs %s -o - | FileCheck %s + + +--- +name: select_freeze_scalar_s32 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: gpr, preferred-register: '', flags: [ ] } + - { id: 1, class: gpr, preferred-register: '', flags: [ ] } +liveins: +fixedStack: +stack: +constants: +body: | + bb.0: + liveins: $edi + + ; CHECK-LABEL: name: select_freeze_scalar_s32 + ; CHECK: [[COPY:%[0-9]+]]:gr32 = COPY $edi + ; CHECK-NOT: G_FREEZE + ; CHECK-NEXT: $eax = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $eax + %0:gpr(s32) = COPY $edi + %1:gpr(s32) = G_FREEZE %0 + $eax = COPY %1(s32) + RET 0, implicit $eax +... + +--- +name: select_freeze_scalar_s64 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: gpr, preferred-register: '', flags: [ ] } + - { id: 1, class: gpr, preferred-register: '', flags: [ ] } +liveins: +fixedStack: +stack: +constants: +body: | + bb.0: + liveins: $rdi + + ; CHECK-LABEL: name: select_freeze_scalar_s64 + ; CHECK: [[COPY:%[0-9]+]]:gr64 = COPY $rdi + ; CHECK-NOT: G_FREEZE + ; CHECK-NEXT: $rax = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $rax + %0:gpr(s64) = COPY $rdi + %1:gpr(s64) = G_FREEZE %0 + $rax = COPY %1(s64) + RET 0, implicit $rax +... + + +--- +name: select_freeze_vec128 +legalized: true +regBankSelected: true +registers: + - { id: 0, class: vecr, preferred-register: '', flags: [ ] } + - { id: 1, class: vecr, preferred-register: '', flags: [ ] } +body: | + bb.0: + liveins: $xmm0 + + ; CHECK-LABEL: name: select_freeze_vec128 + ; CHECK: [[COPY:%[0-9]+]]:vr128 = COPY $xmm0 + ; CHECK-NOT: G_FREEZE + ; CHECK-NEXT: $xmm1 = COPY [[COPY]] + ; CHECK-NEXT: RET 0, implicit $xmm1 + %0:vecr(<4 x s32>) = COPY $xmm0 + %1:vecr(<4 x s32>) = G_FREEZE %0 + $xmm1 = COPY %1(<4 x s32>) + RET 0, implicit $xmm1 +... diff --git a/llvm/test/CodeGen/X86/x86-shrink-wrap-unwind.ll b/llvm/test/CodeGen/X86/x86-shrink-wrap-unwind.ll index 3349d31..b2064b1 100644 --- a/llvm/test/CodeGen/X86/x86-shrink-wrap-unwind.ll +++ b/llvm/test/CodeGen/X86/x86-shrink-wrap-unwind.ll @@ -317,13 +317,13 @@ define void @with_nounwind(i1 %cond) nounwind personality ptr @my_personality { ; CHECK-NEXT: popq %rax ; CHECK-NEXT: retq ; CHECK-NEXT: LBB4_1: ## %throw -; CHECK-NEXT: Ltmp0: +; CHECK-NEXT: Ltmp0: ## EH_LABEL ; CHECK-NEXT: callq _throw_exception -; CHECK-NEXT: Ltmp1: +; CHECK-NEXT: Ltmp1: ## EH_LABEL ; CHECK-NEXT: ## %bb.2: ## %unreachable ; CHECK-NEXT: ud2 ; CHECK-NEXT: LBB4_3: ## %landing -; CHECK-NEXT: Ltmp2: +; CHECK-NEXT: Ltmp2: ## EH_LABEL ; CHECK-NEXT: popq %rax ; CHECK-NEXT: retq ; CHECK-NEXT: Lfunc_end0: @@ -340,12 +340,12 @@ define void @with_nounwind(i1 %cond) nounwind personality ptr @my_personality { ; NOCOMPACTUNWIND-NEXT: retq ; NOCOMPACTUNWIND-NEXT: .LBB4_1: # %throw ; NOCOMPACTUNWIND-NEXT: .cfi_def_cfa_offset 16 -; NOCOMPACTUNWIND-NEXT: .Ltmp0: +; NOCOMPACTUNWIND-NEXT: .Ltmp0: # EH_LABEL ; NOCOMPACTUNWIND-NEXT: callq throw_exception@PLT -; NOCOMPACTUNWIND-NEXT: .Ltmp1: +; NOCOMPACTUNWIND-NEXT: .Ltmp1: # EH_LABEL ; NOCOMPACTUNWIND-NEXT: # %bb.2: # %unreachable ; NOCOMPACTUNWIND-NEXT: .LBB4_3: # %landing -; NOCOMPACTUNWIND-NEXT: .Ltmp2: +; NOCOMPACTUNWIND-NEXT: .Ltmp2: # EH_LABEL ; NOCOMPACTUNWIND-NEXT: popq %rax ; NOCOMPACTUNWIND-NEXT: .cfi_def_cfa_offset 8 ; NOCOMPACTUNWIND-NEXT: retq @@ -379,9 +379,9 @@ define void @with_nounwind_same_succ(i1 %cond) nounwind personality ptr @my_pers ; CHECK-NEXT: ## %bb.1: ## %throw ; CHECK-NEXT: pushq %rax ; CHECK-NEXT: .cfi_def_cfa_offset 16 -; CHECK-NEXT: Ltmp3: +; CHECK-NEXT: Ltmp3: ## EH_LABEL ; CHECK-NEXT: callq _throw_exception -; CHECK-NEXT: Ltmp4: +; CHECK-NEXT: Ltmp4: ## EH_LABEL ; CHECK-NEXT: LBB5_3: ## %fallthrough ; CHECK-NEXT: ## InlineAsm Start ; CHECK-NEXT: nop @@ -390,7 +390,7 @@ define void @with_nounwind_same_succ(i1 %cond) nounwind personality ptr @my_pers ; CHECK-NEXT: LBB5_4: ## %return ; CHECK-NEXT: retq ; CHECK-NEXT: LBB5_2: ## %landing -; CHECK-NEXT: Ltmp5: +; CHECK-NEXT: Ltmp5: ## EH_LABEL ; CHECK-NEXT: jmp LBB5_3 ; CHECK-NEXT: Lfunc_end1: ; @@ -401,9 +401,9 @@ define void @with_nounwind_same_succ(i1 %cond) nounwind personality ptr @my_pers ; NOCOMPACTUNWIND-NEXT: # %bb.1: # %throw ; NOCOMPACTUNWIND-NEXT: pushq %rax ; NOCOMPACTUNWIND-NEXT: .cfi_def_cfa_offset 16 -; NOCOMPACTUNWIND-NEXT: .Ltmp3: +; NOCOMPACTUNWIND-NEXT: .Ltmp3: # EH_LABEL ; NOCOMPACTUNWIND-NEXT: callq throw_exception@PLT -; NOCOMPACTUNWIND-NEXT: .Ltmp4: +; NOCOMPACTUNWIND-NEXT: .Ltmp4: # EH_LABEL ; NOCOMPACTUNWIND-NEXT: .LBB5_3: # %fallthrough ; NOCOMPACTUNWIND-NEXT: #APP ; NOCOMPACTUNWIND-NEXT: nop @@ -414,7 +414,7 @@ define void @with_nounwind_same_succ(i1 %cond) nounwind personality ptr @my_pers ; NOCOMPACTUNWIND-NEXT: retq ; NOCOMPACTUNWIND-NEXT: .LBB5_2: # %landing ; NOCOMPACTUNWIND-NEXT: .cfi_def_cfa_offset 16 -; NOCOMPACTUNWIND-NEXT: .Ltmp5: +; NOCOMPACTUNWIND-NEXT: .Ltmp5: # EH_LABEL ; NOCOMPACTUNWIND-NEXT: jmp .LBB5_3 entry: br i1 %cond, label %throw, label %return diff --git a/llvm/test/ThinLTO/X86/memprof-basic.ll b/llvm/test/ThinLTO/X86/memprof-basic.ll index 0ff0ce0..537e1b8 100644 --- a/llvm/test/ThinLTO/X86/memprof-basic.ll +++ b/llvm/test/ThinLTO/X86/memprof-basic.ll @@ -103,7 +103,9 @@ declare i32 @sleep() define internal ptr @_Z3barv() #0 !dbg !15 { entry: - %call = call ptr @_Znam(i64 0), !memprof !2, !callsite !7 + ;; Use an ambiguous attribute for this allocation, which is now added to such + ;; allocations during matching. It should not affect cloning. + %call = call ptr @_Znam(i64 0) #1, !memprof !2, !callsite !7 ret ptr null } @@ -125,6 +127,7 @@ entry: uselistorder ptr @_Z3foov, { 1, 0 } attributes #0 = { noinline optnone } +attributes #1 = { "memprof"="ambiguous" } !llvm.dbg.cu = !{!13} !llvm.module.flags = !{!20, !21} diff --git a/llvm/test/Transforms/Coroutines/coro-catchswitch-cleanuppad.ll b/llvm/test/Transforms/Coroutines/coro-catchswitch-cleanuppad.ll index d0e7c1c2..e1e1611 100644 --- a/llvm/test/Transforms/Coroutines/coro-catchswitch-cleanuppad.ll +++ b/llvm/test/Transforms/Coroutines/coro-catchswitch-cleanuppad.ll @@ -80,8 +80,8 @@ cleanup2: ; CHECK: cleanup2.corodispatch: ; CHECK: %1 = phi i8 [ 0, %handler2 ], [ 1, %catch.dispatch.2 ] ; CHECK: %2 = cleanuppad within %h1 [] -; CHECK: %switch = icmp ult i8 %1, 1 -; CHECK: br i1 %switch, label %cleanup2.from.handler2, label %cleanup2.from.catch.dispatch.2 +; CHECK: %3 = icmp eq i8 %1, 0 +; CHECK: br i1 %3, label %cleanup2.from.handler2, label %cleanup2.from.catch.dispatch.2 ; CHECK: cleanup2.from.handler2: ; CHECK: %valueB.reload = load i32, ptr %valueB.spill.addr, align 4 diff --git a/llvm/test/Transforms/LoopVectorize/AArch64/pr162009.ll b/llvm/test/Transforms/LoopVectorize/AArch64/pr162009.ll new file mode 100644 index 0000000..6095b24 --- /dev/null +++ b/llvm/test/Transforms/LoopVectorize/AArch64/pr162009.ll @@ -0,0 +1,79 @@ +; NOTE: Assertions have been autogenerated by utils/update_test_checks.py UTC_ARGS: --version 6 +; RUN: opt -passes=loop-vectorize -force-vector-interleave=1 -enable-epilogue-vectorization=false -S < %s | FileCheck %s --check-prefixes=CHECK-NO-PARTIAL-REDUCTION + +target triple = "aarch64" + +define i128 @add_reduc_i32_i128_unsupported(ptr %a, ptr %b) "target-features"="+dotprod" { +; CHECK-NO-PARTIAL-REDUCTION-LABEL: define i128 @add_reduc_i32_i128_unsupported( +; CHECK-NO-PARTIAL-REDUCTION-SAME: ptr [[A:%.*]], ptr [[B:%.*]]) #[[ATTR0:[0-9]+]] { +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[ENTRY:.*:]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: br label %[[VECTOR_PH:.*]] +; CHECK-NO-PARTIAL-REDUCTION: [[VECTOR_PH]]: +; CHECK-NO-PARTIAL-REDUCTION-NEXT: br label %[[VECTOR_BODY:.*]] +; CHECK-NO-PARTIAL-REDUCTION: [[VECTOR_BODY]]: +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[INDEX:%.*]] = phi i64 [ 0, %[[VECTOR_PH]] ], [ [[INDEX_NEXT:%.*]], %[[VECTOR_BODY]] ] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[VEC_PHI:%.*]] = phi <4 x i128> [ zeroinitializer, %[[VECTOR_PH]] ], [ [[TMP7:%.*]], %[[VECTOR_BODY]] ] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP0:%.*]] = getelementptr i32, ptr [[A]], i64 [[INDEX]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[WIDE_LOAD:%.*]] = load <4 x i32>, ptr [[TMP0]], align 1 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP1:%.*]] = zext <4 x i32> [[WIDE_LOAD]] to <4 x i64> +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP2:%.*]] = getelementptr i32, ptr [[B]], i64 [[INDEX]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[WIDE_LOAD1:%.*]] = load <4 x i32>, ptr [[TMP2]], align 1 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP3:%.*]] = zext <4 x i32> [[WIDE_LOAD1]] to <4 x i64> +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP4:%.*]] = mul nuw <4 x i64> [[TMP1]], [[TMP3]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP5:%.*]] = zext <4 x i64> [[TMP4]] to <4 x i128> +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP7]] = add <4 x i128> [[VEC_PHI]], [[TMP5]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[INDEX_NEXT]] = add nuw i64 [[INDEX]], 4 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP6:%.*]] = icmp eq i64 [[INDEX_NEXT]], 4024 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: br i1 [[TMP6]], label %[[MIDDLE_BLOCK:.*]], label %[[VECTOR_BODY]], !llvm.loop [[LOOP0:![0-9]+]] +; CHECK-NO-PARTIAL-REDUCTION: [[MIDDLE_BLOCK]]: +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[TMP8:%.*]] = call i128 @llvm.vector.reduce.add.v4i128(<4 x i128> [[TMP7]]) +; CHECK-NO-PARTIAL-REDUCTION-NEXT: br label %[[SCALAR_PH:.*]] +; CHECK-NO-PARTIAL-REDUCTION: [[SCALAR_PH]]: +; CHECK-NO-PARTIAL-REDUCTION-NEXT: br label %[[FOR_BODY:.*]] +; CHECK-NO-PARTIAL-REDUCTION: [[FOR_BODY]]: +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[IV:%.*]] = phi i64 [ 4024, %[[SCALAR_PH]] ], [ [[IV_NEXT:%.*]], %[[FOR_BODY]] ] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[ACCUM:%.*]] = phi i128 [ [[TMP8]], %[[SCALAR_PH]] ], [ [[ADD:%.*]], %[[FOR_BODY]] ] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[GEP_A:%.*]] = getelementptr i32, ptr [[A]], i64 [[IV]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[LOAD_A:%.*]] = load i32, ptr [[GEP_A]], align 1 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[EXT_A:%.*]] = zext i32 [[LOAD_A]] to i64 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[GEP_B:%.*]] = getelementptr i32, ptr [[B]], i64 [[IV]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[LOAD_B:%.*]] = load i32, ptr [[GEP_B]], align 1 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[EXT_B:%.*]] = zext i32 [[LOAD_B]] to i64 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[MUL:%.*]] = mul nuw i64 [[EXT_A]], [[EXT_B]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[MUL_ZEXT:%.*]] = zext i64 [[MUL]] to i128 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[ADD]] = add i128 [[ACCUM]], [[MUL_ZEXT]] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[IV_NEXT]] = add i64 [[IV]], 1 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[EXITCOND_NOT:%.*]] = icmp eq i64 [[IV_NEXT]], 4025 +; CHECK-NO-PARTIAL-REDUCTION-NEXT: br i1 [[EXITCOND_NOT]], label %[[FOR_EXIT:.*]], label %[[FOR_BODY]], !llvm.loop [[LOOP3:![0-9]+]] +; CHECK-NO-PARTIAL-REDUCTION: [[FOR_EXIT]]: +; CHECK-NO-PARTIAL-REDUCTION-NEXT: [[ADD_LCSSA:%.*]] = phi i128 [ [[ADD]], %[[FOR_BODY]] ] +; CHECK-NO-PARTIAL-REDUCTION-NEXT: ret i128 [[ADD_LCSSA]] +; +entry: + br label %for.body + +for.body: + %iv = phi i64 [ 0, %entry ], [ %iv.next, %for.body ] + %accum = phi i128 [ 0, %entry ], [ %add, %for.body ] + %gep.a = getelementptr i32, ptr %a, i64 %iv + %load.a = load i32, ptr %gep.a, align 1 + %ext.a = zext i32 %load.a to i64 + %gep.b = getelementptr i32, ptr %b, i64 %iv + %load.b = load i32, ptr %gep.b, align 1 + %ext.b = zext i32 %load.b to i64 + %mul = mul nuw i64 %ext.a, %ext.b + %mul.zext = zext i64 %mul to i128 + %add = add i128 %accum, %mul.zext + %iv.next = add i64 %iv, 1 + %exitcond.not = icmp eq i64 %iv.next, 4025 + br i1 %exitcond.not, label %for.exit, label %for.body + +for.exit: + ret i128 %add +} +;. +; CHECK-NO-PARTIAL-REDUCTION: [[LOOP0]] = distinct !{[[LOOP0]], [[META1:![0-9]+]], [[META2:![0-9]+]]} +; CHECK-NO-PARTIAL-REDUCTION: [[META1]] = !{!"llvm.loop.isvectorized", i32 1} +; CHECK-NO-PARTIAL-REDUCTION: [[META2]] = !{!"llvm.loop.unroll.runtime.disable"} +; CHECK-NO-PARTIAL-REDUCTION: [[LOOP3]] = distinct !{[[LOOP3]], [[META2]], [[META1]]} +;. diff --git a/llvm/test/Transforms/PGOProfile/memprof.ll b/llvm/test/Transforms/PGOProfile/memprof.ll index c69d031..f6a89a8 100644 --- a/llvm/test/Transforms/PGOProfile/memprof.ll +++ b/llvm/test/Transforms/PGOProfile/memprof.ll @@ -38,7 +38,7 @@ ; ALL-NOT: no profile data available for function ;; Using a memprof-only profile for memprof-use should only give memprof metadata -; RUN: opt < %s -passes='memprof-use<profile-filename=%t.memprofdata>' -pgo-warn-missing-function -S -memprof-print-match-info -stats 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,MEMPROFONLY,MEMPROFMATCHINFO,MEMPROFSTATS +; RUN: opt < %s -passes='memprof-use<profile-filename=%t.memprofdata>' -pgo-warn-missing-function -S -memprof-print-match-info -stats 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,MEMPROFONLY,MEMPROFMATCHINFO,MEMPROFSTATS,AMBIG ; There should not be any PGO metadata ; MEMPROFONLY-NOT: !prof @@ -51,10 +51,10 @@ ;; Test the same thing but by passing the memory profile through to a default ;; pipeline via -memory-profile-file=, which should cause the necessary field ;; of the PGOOptions structure to be populated with the profile filename. -; RUN: opt < %s -passes='default<O2>' -memory-profile-file=%t.memprofdata -pgo-warn-missing-function -S 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,MEMPROFONLY +; RUN: opt < %s -passes='default<O2>' -memory-profile-file=%t.memprofdata -pgo-warn-missing-function -S 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,MEMPROFONLY,AMBIG ;; Using a pgo+memprof profile for memprof-use should only give memprof metadata -; RUN: opt < %s -passes='memprof-use<profile-filename=%t.pgomemprofdata>' -pgo-warn-missing-function -S 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,MEMPROFONLY +; RUN: opt < %s -passes='memprof-use<profile-filename=%t.pgomemprofdata>' -pgo-warn-missing-function -S 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,MEMPROFONLY,AMBIG ;; Using a pgo-only profile for memprof-use should give an error ; RUN: not opt < %s -passes='memprof-use<profile-filename=%t.pgoprofdata>' -S 2>&1 | FileCheck %s --check-prefixes=MEMPROFWITHPGOONLY @@ -72,7 +72,7 @@ ;; Using a pgo+memprof profile for both memprof-use and pgo-instr-use should ;; give both memprof and pgo metadata. -; RUN: opt < %s -passes='pgo-instr-use,memprof-use<profile-filename=%t.pgomemprofdata>' -pgo-test-profile-file=%t.pgomemprofdata -pgo-warn-missing-function -S 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,PGO +; RUN: opt < %s -passes='pgo-instr-use,memprof-use<profile-filename=%t.pgomemprofdata>' -pgo-test-profile-file=%t.pgomemprofdata -pgo-warn-missing-function -S 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,PGO,AMBIG ;; Check that the total sizes are reported if requested. A message should be ;; emitted for the pruned context. Also check that remarks are emitted for the @@ -108,7 +108,11 @@ ;; However, with the same threshold, but hot hints not enabled, it should be ;; notcold again. -; RUN: opt < %s -passes='memprof-use<profile-filename=%t.memprofdata>' -pgo-warn-missing-function -S -memprof-min-ave-lifetime-access-density-hot-threshold=0 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL +; RUN: opt < %s -passes='memprof-use<profile-filename=%t.memprofdata>' -pgo-warn-missing-function -S -memprof-min-ave-lifetime-access-density-hot-threshold=0 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,AMBIG + +;; Test that we don't get an ambiguous memprof attribute when +;; -memprof-ambiguous-attributes is disabled. +; RUN: opt < %s -passes='memprof-use<profile-filename=%t.memprofdata>' -pgo-warn-missing-function -S -memprof-ambiguous-attributes=false 2>&1 | FileCheck %s --check-prefixes=MEMPROF,ALL,NOAMBIG ; MEMPROFMATCHINFO: MemProf notcold context with id 1093248920606587996 has total profiled size 10 is matched with 1 frames ; MEMPROFMATCHINFO: MemProf notcold context with id 5725971306423925017 has total profiled size 10 is matched with 1 frames @@ -140,7 +144,7 @@ target triple = "x86_64-unknown-linux-gnu" ; PGO: !prof define dso_local noundef ptr @_Z3foov() #0 !dbg !10 { entry: - ; MEMPROF: call {{.*}} @_Znam{{.*}} !memprof ![[M1:[0-9]+]], !callsite ![[C1:[0-9]+]] + ; MEMPROF: call {{.*}} @_Znam{{.*}} #[[A0:[0-9]+]]{{.*}} !memprof ![[M1:[0-9]+]], !callsite ![[C1:[0-9]+]] ; MEMPROFNOCOLINFO: call {{.*}} @_Znam{{.*}} !memprof ![[M1:[0-9]+]], !callsite ![[C1:[0-9]+]] %call = call noalias noundef nonnull ptr @_Znam(i64 noundef 10) #6, !dbg !13 ret ptr %call, !dbg !14 @@ -364,6 +368,9 @@ for.end: ; preds = %for.cond ret i32 0, !dbg !103 } +;; We optionally apply an ambiguous memprof attribute to ambiguous allocations +; AMBIG: #[[A0]] = { builtin allocsize(0) "memprof"="ambiguous" } +; NOAMBIG: #[[A0]] = { builtin allocsize(0) } ; MEMPROF: #[[A1]] = { builtin allocsize(0) "memprof"="notcold" } ; MEMPROF: #[[A2]] = { builtin allocsize(0) "memprof"="cold" } ; MEMPROF: ![[M1]] = !{![[MIB1:[0-9]+]], ![[MIB2:[0-9]+]], ![[MIB3:[0-9]+]], ![[MIB4:[0-9]+]]} diff --git a/llvm/test/Transforms/SCCP/relax-range-checks.ll b/llvm/test/Transforms/SCCP/relax-range-checks.ll index 90722f3..34e4813 100644 --- a/llvm/test/Transforms/SCCP/relax-range-checks.ll +++ b/llvm/test/Transforms/SCCP/relax-range-checks.ll @@ -89,4 +89,28 @@ define i1 @relax_range_check_multiuse(i8 range(i8 0, 5) %x) { ret i1 %ret } +define i1 @range_check_to_icmp_eq1(i32 range(i32 0, 4) %x) { +; CHECK-LABEL: define i1 @range_check_to_icmp_eq1( +; CHECK-SAME: i32 range(i32 0, 4) [[X:%.*]]) { +; CHECK-NEXT: [[OFF:%.*]] = add nsw i32 [[X]], -3 +; CHECK-NEXT: [[TMP1:%.*]] = icmp eq i32 [[X]], 3 +; CHECK-NEXT: ret i1 [[TMP1]] +; + %off = add nsw i32 %x, -3 + %cmp = icmp ult i32 %off, 2 + ret i1 %cmp +} + +define i1 @range_check_to_icmp_eq2(i32 range(i32 -1, 2) %x) { +; CHECK-LABEL: define i1 @range_check_to_icmp_eq2( +; CHECK-SAME: i32 range(i32 -1, 2) [[X:%.*]]) { +; CHECK-NEXT: [[OFF:%.*]] = add nsw i32 [[X]], -1 +; CHECK-NEXT: [[CMP:%.*]] = icmp eq i32 [[X]], 1 +; CHECK-NEXT: ret i1 [[CMP]] +; + %off = add nsw i32 %x, -1 + %cmp = icmp ult i32 %off, -2 + ret i1 %cmp +} + declare void @use(i8) diff --git a/llvm/test/Transforms/SLPVectorizer/RISCV/strided-loads-with-external-indices.ll b/llvm/test/Transforms/SLPVectorizer/RISCV/strided-loads-with-external-indices.ll index 655db54..a079203 100644 --- a/llvm/test/Transforms/SLPVectorizer/RISCV/strided-loads-with-external-indices.ll +++ b/llvm/test/Transforms/SLPVectorizer/RISCV/strided-loads-with-external-indices.ll @@ -10,14 +10,10 @@ define void @test() { ; CHECK-NEXT: [[SUB4_I_I65_US:%.*]] = or i64 0, 1 ; CHECK-NEXT: br label [[BODY:%.*]] ; CHECK: body: -; CHECK-NEXT: [[ADD_I_I62_US:%.*]] = shl i64 0, 0 -; CHECK-NEXT: [[TMP0:%.*]] = insertelement <2 x i64> <i64 poison, i64 1>, i64 [[ADD_I_I62_US]], i32 0 -; CHECK-NEXT: [[TMP1:%.*]] = or <2 x i64> zeroinitializer, [[TMP0]] -; CHECK-NEXT: [[TMP2:%.*]] = getelementptr [[CLASS_A:%.*]], <2 x ptr> zeroinitializer, <2 x i64> [[TMP1]] -; CHECK-NEXT: [[TMP3:%.*]] = call <2 x i32> @llvm.masked.gather.v2i32.v2p0(<2 x ptr> [[TMP2]], i32 4, <2 x i1> splat (i1 true), <2 x i32> poison) -; CHECK-NEXT: [[TMP4:%.*]] = extractelement <2 x i32> [[TMP3]], i32 0 -; CHECK-NEXT: [[TMP5:%.*]] = extractelement <2 x i32> [[TMP3]], i32 1 -; CHECK-NEXT: [[CMP_I_I_I_I67_US:%.*]] = icmp slt i32 [[TMP4]], [[TMP5]] +; CHECK-NEXT: [[TMP0:%.*]] = call <2 x i32> @llvm.masked.gather.v2i32.v2p0(<2 x ptr> getelementptr ([[CLASS_A:%.*]], <2 x ptr> zeroinitializer, <2 x i64> <i64 0, i64 1>), i32 4, <2 x i1> splat (i1 true), <2 x i32> poison) +; CHECK-NEXT: [[TMP1:%.*]] = extractelement <2 x i32> [[TMP0]], i32 0 +; CHECK-NEXT: [[TMP2:%.*]] = extractelement <2 x i32> [[TMP0]], i32 1 +; CHECK-NEXT: [[CMP_I_I_I_I67_US:%.*]] = icmp slt i32 [[TMP1]], [[TMP2]] ; CHECK-NEXT: [[SPEC_SELECT_I_I68_US:%.*]] = select i1 false, i64 [[SUB4_I_I65_US]], i64 0 ; CHECK-NEXT: br label [[BODY]] ; diff --git a/llvm/test/Transforms/SLPVectorizer/X86/ext-used-scalar-different-bitwidth.ll b/llvm/test/Transforms/SLPVectorizer/X86/ext-used-scalar-different-bitwidth.ll index 7758596..87f2cca 100644 --- a/llvm/test/Transforms/SLPVectorizer/X86/ext-used-scalar-different-bitwidth.ll +++ b/llvm/test/Transforms/SLPVectorizer/X86/ext-used-scalar-different-bitwidth.ll @@ -8,8 +8,8 @@ define i32 @test() { ; CHECK-NEXT: [[ENTRY:.*:]] ; CHECK-NEXT: store i32 152, ptr @f, align 4 ; CHECK-NEXT: [[AGG_TMP_SROA_0_0_COPYLOAD_I:%.*]] = load i32, ptr @f, align 4 -; CHECK-NEXT: [[ADD_I_I:%.*]] = shl i32 [[AGG_TMP_SROA_0_0_COPYLOAD_I]], 24 -; CHECK-NEXT: [[TMP0:%.*]] = insertelement <8 x i32> <i32 poison, i32 83886080, i32 83886080, i32 83886080, i32 83886080, i32 83886080, i32 83886080, i32 83886080>, i32 [[ADD_I_I]], i32 0 +; CHECK-NEXT: [[TMP3:%.*]] = insertelement <8 x i32> <i32 poison, i32 83886080, i32 83886080, i32 83886080, i32 83886080, i32 83886080, i32 83886080, i32 83886080>, i32 [[AGG_TMP_SROA_0_0_COPYLOAD_I]], i32 0 +; CHECK-NEXT: [[TMP0:%.*]] = shl <8 x i32> [[TMP3]], <i32 24, i32 0, i32 0, i32 0, i32 0, i32 0, i32 0, i32 0> ; CHECK-NEXT: [[TMP1:%.*]] = add <8 x i32> <i32 83886080, i32 0, i32 0, i32 0, i32 0, i32 0, i32 0, i32 0>, [[TMP0]] ; CHECK-NEXT: [[TMP2:%.*]] = ashr <8 x i32> [[TMP1]], splat (i32 24) ; CHECK-NEXT: [[TMP5:%.*]] = and <8 x i32> [[TMP2]], <i32 66440127, i32 1, i32 1, i32 1, i32 1, i32 1, i32 1, i32 1> diff --git a/llvm/test/Transforms/SLPVectorizer/X86/vect_copyable_in_binops.ll b/llvm/test/Transforms/SLPVectorizer/X86/vect_copyable_in_binops.ll index 75aec45..3e0a374 100644 --- a/llvm/test/Transforms/SLPVectorizer/X86/vect_copyable_in_binops.ll +++ b/llvm/test/Transforms/SLPVectorizer/X86/vect_copyable_in_binops.ll @@ -247,32 +247,12 @@ entry: } define void @shl0(ptr noalias %dst, ptr noalias %src) { -; NON-POW2-LABEL: @shl0( -; NON-POW2-NEXT: entry: -; NON-POW2-NEXT: [[INCDEC_PTR:%.*]] = getelementptr inbounds i32, ptr [[SRC:%.*]], i64 1 -; NON-POW2-NEXT: [[TMP0:%.*]] = load i32, ptr [[SRC]], align 4 -; NON-POW2-NEXT: [[INCDEC_PTR1:%.*]] = getelementptr inbounds i32, ptr [[DST:%.*]], i64 1 -; NON-POW2-NEXT: store i32 [[TMP0]], ptr [[DST]], align 4 -; NON-POW2-NEXT: [[TMP1:%.*]] = load <3 x i32>, ptr [[INCDEC_PTR]], align 4 -; NON-POW2-NEXT: [[TMP2:%.*]] = shl <3 x i32> [[TMP1]], <i32 1, i32 2, i32 3> -; NON-POW2-NEXT: store <3 x i32> [[TMP2]], ptr [[INCDEC_PTR1]], align 4 -; NON-POW2-NEXT: ret void -; -; POW2-ONLY-LABEL: @shl0( -; POW2-ONLY-NEXT: entry: -; POW2-ONLY-NEXT: [[INCDEC_PTR:%.*]] = getelementptr inbounds i32, ptr [[SRC:%.*]], i64 1 -; POW2-ONLY-NEXT: [[TMP0:%.*]] = load i32, ptr [[SRC]], align 4 -; POW2-ONLY-NEXT: [[INCDEC_PTR1:%.*]] = getelementptr inbounds i32, ptr [[DST:%.*]], i64 1 -; POW2-ONLY-NEXT: store i32 [[TMP0]], ptr [[DST]], align 4 -; POW2-ONLY-NEXT: [[INCDEC_PTR4:%.*]] = getelementptr inbounds i32, ptr [[SRC]], i64 3 -; POW2-ONLY-NEXT: [[INCDEC_PTR6:%.*]] = getelementptr inbounds i32, ptr [[DST]], i64 3 -; POW2-ONLY-NEXT: [[TMP1:%.*]] = load <2 x i32>, ptr [[INCDEC_PTR]], align 4 -; POW2-ONLY-NEXT: [[TMP2:%.*]] = shl <2 x i32> [[TMP1]], <i32 1, i32 2> -; POW2-ONLY-NEXT: store <2 x i32> [[TMP2]], ptr [[INCDEC_PTR1]], align 4 -; POW2-ONLY-NEXT: [[TMP3:%.*]] = load i32, ptr [[INCDEC_PTR4]], align 4 -; POW2-ONLY-NEXT: [[SHL8:%.*]] = shl i32 [[TMP3]], 3 -; POW2-ONLY-NEXT: store i32 [[SHL8]], ptr [[INCDEC_PTR6]], align 4 -; POW2-ONLY-NEXT: ret void +; CHECK-LABEL: @shl0( +; CHECK-NEXT: entry: +; CHECK-NEXT: [[TMP0:%.*]] = load <4 x i32>, ptr [[SRC:%.*]], align 4 +; CHECK-NEXT: [[TMP1:%.*]] = shl <4 x i32> [[TMP0]], <i32 0, i32 1, i32 2, i32 3> +; CHECK-NEXT: store <4 x i32> [[TMP1]], ptr [[DST:%.*]], align 4 +; CHECK-NEXT: ret void ; entry: %incdec.ptr = getelementptr inbounds i32, ptr %src, i64 1 diff --git a/llvm/test/Transforms/SLPVectorizer/bool-logical-op-reduction-with-poison.ll b/llvm/test/Transforms/SLPVectorizer/bool-logical-op-reduction-with-poison.ll index a5b1e9b..769b360 100644 --- a/llvm/test/Transforms/SLPVectorizer/bool-logical-op-reduction-with-poison.ll +++ b/llvm/test/Transforms/SLPVectorizer/bool-logical-op-reduction-with-poison.ll @@ -1,25 +1,44 @@ ; NOTE: Assertions have been autogenerated by utils/update_test_checks.py UTC_ARGS: --version 3 -; RUN: %if x86-registered-target %{ opt -S --passes=slp-vectorizer < %s -mtriple=x86_64-unknown-linux-gnu | FileCheck %s %} -; RUN: %if aarch64-registered-target %{ opt -S --passes=slp-vectorizer < %s -mtriple=aarch64-unknown-linux-gnu | FileCheck %s %} +; RUN: %if x86-registered-target %{ opt -S --passes=slp-vectorizer < %s -mtriple=x86_64-unknown-linux-gnu | FileCheck %s --check-prefix=X86 %} +; RUN: %if aarch64-registered-target %{ opt -S --passes=slp-vectorizer < %s -mtriple=aarch64-unknown-linux-gnu | FileCheck %s --check-prefix=AARCH64 %} define i1 @test(i32 %0, i32 %1, i32 %p) { -; CHECK-LABEL: define i1 @test( -; CHECK-SAME: i32 [[TMP0:%.*]], i32 [[TMP1:%.*]], i32 [[P:%.*]]) { -; CHECK-NEXT: entry: -; CHECK-NEXT: [[CMP1:%.*]] = icmp sgt i32 [[TMP0]], 0 -; CHECK-NEXT: [[TMP2:%.*]] = insertelement <4 x i32> poison, i32 [[TMP1]], i32 0 -; CHECK-NEXT: [[TMP3:%.*]] = shufflevector <4 x i32> [[TMP2]], <4 x i32> poison, <4 x i32> zeroinitializer -; CHECK-NEXT: [[TMP4:%.*]] = shl <4 x i32> zeroinitializer, [[TMP3]] -; CHECK-NEXT: [[TMP5:%.*]] = icmp slt <4 x i32> [[TMP4]], zeroinitializer -; CHECK-NEXT: [[CMP6:%.*]] = icmp slt i32 0, [[P]] -; CHECK-NEXT: [[TMP6:%.*]] = freeze <4 x i1> [[TMP5]] -; CHECK-NEXT: [[TMP7:%.*]] = call i1 @llvm.vector.reduce.or.v4i1(<4 x i1> [[TMP6]]) -; CHECK-NEXT: [[OP_RDX:%.*]] = select i1 [[TMP7]], i1 true, i1 [[CMP6]] -; CHECK-NEXT: [[OP_RDX1:%.*]] = select i1 [[CMP1]], i1 true, i1 [[CMP1]] -; CHECK-NEXT: [[TMP8:%.*]] = freeze i1 [[OP_RDX]] -; CHECK-NEXT: [[OP_RDX2:%.*]] = select i1 [[TMP8]], i1 true, i1 [[OP_RDX1]] -; CHECK-NEXT: ret i1 [[OP_RDX2]] +; X86-LABEL: define i1 @test( +; X86-SAME: i32 [[TMP0:%.*]], i32 [[TMP1:%.*]], i32 [[P:%.*]]) { +; X86-NEXT: entry: +; X86-NEXT: [[CMP1:%.*]] = icmp sgt i32 [[TMP0]], 0 +; X86-NEXT: [[TMP2:%.*]] = insertelement <4 x i32> poison, i32 [[TMP1]], i32 0 +; X86-NEXT: [[TMP3:%.*]] = shufflevector <4 x i32> [[TMP2]], <4 x i32> poison, <4 x i32> zeroinitializer +; X86-NEXT: [[TMP4:%.*]] = shl <4 x i32> zeroinitializer, [[TMP3]] +; X86-NEXT: [[TMP5:%.*]] = icmp slt <4 x i32> [[TMP4]], zeroinitializer +; X86-NEXT: [[CMP6:%.*]] = icmp slt i32 0, [[P]] +; X86-NEXT: [[TMP6:%.*]] = freeze <4 x i1> [[TMP5]] +; X86-NEXT: [[TMP7:%.*]] = call i1 @llvm.vector.reduce.or.v4i1(<4 x i1> [[TMP6]]) +; X86-NEXT: [[OP_RDX:%.*]] = select i1 [[TMP7]], i1 true, i1 [[CMP6]] +; X86-NEXT: [[OP_RDX1:%.*]] = select i1 [[CMP1]], i1 true, i1 [[CMP1]] +; X86-NEXT: [[TMP8:%.*]] = freeze i1 [[OP_RDX]] +; X86-NEXT: [[OP_RDX2:%.*]] = select i1 [[TMP8]], i1 true, i1 [[OP_RDX1]] +; X86-NEXT: ret i1 [[OP_RDX2]] +; +; AARCH64-LABEL: define i1 @test( +; AARCH64-SAME: i32 [[TMP0:%.*]], i32 [[TMP1:%.*]], i32 [[P:%.*]]) { +; AARCH64-NEXT: entry: +; AARCH64-NEXT: [[CMP1:%.*]] = icmp sgt i32 [[TMP0]], 0 +; AARCH64-NEXT: [[SHL4:%.*]] = shl i32 0, [[TMP1]] +; AARCH64-NEXT: [[CMP5:%.*]] = icmp slt i32 [[SHL4]], 0 +; AARCH64-NEXT: [[TMP2:%.*]] = insertelement <4 x i32> <i32 0, i32 poison, i32 poison, i32 poison>, i32 [[TMP1]], i32 1 +; AARCH64-NEXT: [[TMP3:%.*]] = shufflevector <4 x i32> [[TMP2]], <4 x i32> poison, <4 x i32> <i32 0, i32 1, i32 1, i32 1> +; AARCH64-NEXT: [[TMP4:%.*]] = shl <4 x i32> zeroinitializer, [[TMP3]] +; AARCH64-NEXT: [[TMP5:%.*]] = insertelement <4 x i32> <i32 poison, i32 0, i32 0, i32 0>, i32 [[P]], i32 0 +; AARCH64-NEXT: [[TMP6:%.*]] = icmp slt <4 x i32> [[TMP4]], [[TMP5]] +; AARCH64-NEXT: [[TMP7:%.*]] = freeze <4 x i1> [[TMP6]] +; AARCH64-NEXT: [[TMP8:%.*]] = call i1 @llvm.vector.reduce.or.v4i1(<4 x i1> [[TMP7]]) +; AARCH64-NEXT: [[OP_RDX:%.*]] = select i1 [[TMP8]], i1 true, i1 [[CMP5]] +; AARCH64-NEXT: [[OP_RDX1:%.*]] = select i1 [[CMP1]], i1 true, i1 [[CMP1]] +; AARCH64-NEXT: [[TMP9:%.*]] = freeze i1 [[OP_RDX]] +; AARCH64-NEXT: [[OP_RDX2:%.*]] = select i1 [[TMP9]], i1 true, i1 [[OP_RDX1]] +; AARCH64-NEXT: ret i1 [[OP_RDX2]] ; entry: %cmp1 = icmp sgt i32 %0, 0 diff --git a/llvm/test/Transforms/SimplifyCFG/switch-dead-default.ll b/llvm/test/Transforms/SimplifyCFG/switch-dead-default.ll index 4a457cc..a0e29dd 100644 --- a/llvm/test/Transforms/SimplifyCFG/switch-dead-default.ll +++ b/llvm/test/Transforms/SimplifyCFG/switch-dead-default.ll @@ -7,8 +7,7 @@ declare void @foo(i32) define void @test(i1 %a) { ; CHECK-LABEL: define void @test( ; CHECK-SAME: i1 [[A:%.*]]) { -; CHECK-NEXT: [[A_OFF:%.*]] = add i1 [[A]], true -; CHECK-NEXT: [[SWITCH:%.*]] = icmp ult i1 [[A_OFF]], true +; CHECK-NEXT: [[SWITCH:%.*]] = icmp eq i1 [[A]], true ; CHECK-NEXT: br i1 [[SWITCH]], label [[TRUE:%.*]], label [[FALSE:%.*]] ; CHECK: common.ret: ; CHECK-NEXT: ret void @@ -209,8 +208,7 @@ define void @test5(i8 %a) { ; CHECK-SAME: i8 [[A:%.*]]) { ; CHECK-NEXT: [[CMP:%.*]] = icmp ult i8 [[A]], 2 ; CHECK-NEXT: call void @llvm.assume(i1 [[CMP]]) -; CHECK-NEXT: [[A_OFF:%.*]] = add i8 [[A]], -1 -; CHECK-NEXT: [[SWITCH:%.*]] = icmp ult i8 [[A_OFF]], 1 +; CHECK-NEXT: [[SWITCH:%.*]] = icmp eq i8 [[A]], 1 ; CHECK-NEXT: br i1 [[SWITCH]], label [[TRUE:%.*]], label [[FALSE:%.*]] ; CHECK: common.ret: ; CHECK-NEXT: ret void @@ -243,8 +241,7 @@ define void @test6(i8 %a) { ; CHECK-NEXT: [[AND:%.*]] = and i8 [[A]], -2 ; CHECK-NEXT: [[CMP:%.*]] = icmp eq i8 [[AND]], -2 ; CHECK-NEXT: call void @llvm.assume(i1 [[CMP]]) -; CHECK-NEXT: [[A_OFF:%.*]] = add i8 [[A]], 1 -; CHECK-NEXT: [[SWITCH:%.*]] = icmp ult i8 [[A_OFF]], 1 +; CHECK-NEXT: [[SWITCH:%.*]] = icmp eq i8 [[A]], -1 ; CHECK-NEXT: br i1 [[SWITCH]], label [[TRUE:%.*]], label [[FALSE:%.*]] ; CHECK: common.ret: ; CHECK-NEXT: ret void @@ -279,8 +276,7 @@ define void @test7(i8 %a) { ; CHECK-NEXT: [[AND:%.*]] = and i8 [[A]], -2 ; CHECK-NEXT: [[CMP:%.*]] = icmp eq i8 [[AND]], -2 ; CHECK-NEXT: call void @llvm.assume(i1 [[CMP]]) -; CHECK-NEXT: [[A_OFF:%.*]] = add i8 [[A]], 1 -; CHECK-NEXT: [[SWITCH:%.*]] = icmp ult i8 [[A_OFF]], 1 +; CHECK-NEXT: [[SWITCH:%.*]] = icmp eq i8 [[A]], -1 ; CHECK-NEXT: br i1 [[SWITCH]], label [[TRUE:%.*]], label [[FALSE:%.*]] ; CHECK: common.ret: ; CHECK-NEXT: ret void diff --git a/llvm/test/Transforms/SimplifyCFG/switch-range-to-icmp.ll b/llvm/test/Transforms/SimplifyCFG/switch-range-to-icmp.ll index 8f2ae2d..0fc3c19 100644 --- a/llvm/test/Transforms/SimplifyCFG/switch-range-to-icmp.ll +++ b/llvm/test/Transforms/SimplifyCFG/switch-range-to-icmp.ll @@ -188,4 +188,217 @@ exit: ret void } +define i32 @wrapping_known_range(i8 range(i8 0, 6) %arg) { +; CHECK-LABEL: @wrapping_known_range( +; CHECK-NEXT: [[ARG_OFF:%.*]] = add i8 [[ARG:%.*]], -1 +; CHECK-NEXT: [[SWITCH:%.*]] = icmp ult i8 [[ARG_OFF]], 3 +; CHECK-NEXT: br i1 [[SWITCH]], label [[ELSE:%.*]], label [[IF:%.*]] +; CHECK: common.ret: +; CHECK-NEXT: [[COMMON_RET_OP:%.*]] = phi i32 [ [[I0:%.*]], [[IF]] ], [ [[I1:%.*]], [[ELSE]] ] +; CHECK-NEXT: ret i32 [[COMMON_RET_OP]] +; CHECK: if: +; CHECK-NEXT: [[I0]] = call i32 @f(i32 0) +; CHECK-NEXT: br label [[COMMON_RET:%.*]] +; CHECK: else: +; CHECK-NEXT: [[I1]] = call i32 @f(i32 1) +; CHECK-NEXT: br label [[COMMON_RET]] +; + switch i8 %arg, label %else [ + i8 0, label %if + i8 4, label %if + i8 5, label %if + ] + +if: + %i0 = call i32 @f(i32 0) + ret i32 %i0 + +else: + %i1 = call i32 @f(i32 1) + ret i32 %i1 +} + +define i32 @wrapping_known_range_2(i8 range(i8 0, 6) %arg) { +; CHECK-LABEL: @wrapping_known_range_2( +; CHECK-NEXT: [[SWITCH:%.*]] = icmp eq i8 [[ARG:%.*]], 1 +; CHECK-NEXT: br i1 [[SWITCH]], label [[ELSE:%.*]], label [[IF:%.*]] +; CHECK: common.ret: +; CHECK-NEXT: [[COMMON_RET_OP:%.*]] = phi i32 [ [[I0:%.*]], [[IF]] ], [ [[I1:%.*]], [[ELSE]] ] +; CHECK-NEXT: ret i32 [[COMMON_RET_OP]] +; CHECK: if: +; CHECK-NEXT: [[I0]] = call i32 @f(i32 0) +; CHECK-NEXT: br label [[COMMON_RET:%.*]] +; CHECK: else: +; CHECK-NEXT: [[I1]] = call i32 @f(i32 1) +; CHECK-NEXT: br label [[COMMON_RET]] +; + switch i8 %arg, label %else [ + i8 0, label %if + i8 2, label %if + i8 3, label %if + i8 4, label %if + i8 5, label %if + ] + +if: + %i0 = call i32 @f(i32 0) + ret i32 %i0 + +else: + %i1 = call i32 @f(i32 1) + ret i32 %i1 +} + +define i32 @wrapping_range(i8 %arg) { +; CHECK-LABEL: @wrapping_range( +; CHECK-NEXT: [[ARG_OFF:%.*]] = add i8 [[ARG:%.*]], -1 +; CHECK-NEXT: [[SWITCH:%.*]] = icmp ult i8 [[ARG_OFF]], -4 +; CHECK-NEXT: br i1 [[SWITCH]], label [[ELSE:%.*]], label [[IF:%.*]] +; CHECK: common.ret: +; CHECK-NEXT: [[COMMON_RET_OP:%.*]] = phi i32 [ [[I0:%.*]], [[IF]] ], [ [[I1:%.*]], [[ELSE]] ] +; CHECK-NEXT: ret i32 [[COMMON_RET_OP]] +; CHECK: if: +; CHECK-NEXT: [[I0]] = call i32 @f(i32 0) +; CHECK-NEXT: br label [[COMMON_RET:%.*]] +; CHECK: else: +; CHECK-NEXT: [[I1]] = call i32 @f(i32 1) +; CHECK-NEXT: br label [[COMMON_RET]] +; + switch i8 %arg, label %else [ + i8 0, label %if + i8 -3, label %if + i8 -2, label %if + i8 -1, label %if + ] + +if: + %i0 = call i32 @f(i32 0) + ret i32 %i0 + +else: + %i1 = call i32 @f(i32 1) + ret i32 %i1 +} + +define i8 @wrapping_range_phi(i8 %arg) { +; CHECK-LABEL: @wrapping_range_phi( +; CHECK-NEXT: entry: +; CHECK-NEXT: [[ARG_OFF:%.*]] = add i8 [[ARG:%.*]], -1 +; CHECK-NEXT: [[SWITCH:%.*]] = icmp ult i8 [[ARG_OFF]], -2 +; CHECK-NEXT: [[SPEC_SELECT:%.*]] = select i1 [[SWITCH]], i8 0, i8 1 +; CHECK-NEXT: ret i8 [[SPEC_SELECT]] +; +entry: + switch i8 %arg, label %else [ + i8 0, label %if + i8 -1, label %if + ] + +if: + %i = phi i8 [ 0, %else ], [ 1, %entry ], [ 1, %entry ] + ret i8 %i + +else: + br label %if +} + +define i32 @no_continuous_wrapping_range(i8 %arg) { +; CHECK-LABEL: @no_continuous_wrapping_range( +; CHECK-NEXT: switch i8 [[ARG:%.*]], label [[ELSE:%.*]] [ +; CHECK-NEXT: i8 0, label [[IF:%.*]] +; CHECK-NEXT: i8 -3, label [[IF]] +; CHECK-NEXT: i8 -1, label [[IF]] +; CHECK-NEXT: ] +; CHECK: common.ret: +; CHECK-NEXT: [[COMMON_RET_OP:%.*]] = phi i32 [ [[I0:%.*]], [[IF]] ], [ [[I1:%.*]], [[ELSE]] ] +; CHECK-NEXT: ret i32 [[COMMON_RET_OP]] +; CHECK: if: +; CHECK-NEXT: [[I0]] = call i32 @f(i32 0) +; CHECK-NEXT: br label [[COMMON_RET:%.*]] +; CHECK: else: +; CHECK-NEXT: [[I1]] = call i32 @f(i32 1) +; CHECK-NEXT: br label [[COMMON_RET]] +; + switch i8 %arg, label %else [ + i8 0, label %if + i8 -3, label %if + i8 -1, label %if + ] + +if: + %i0 = call i32 @f(i32 0) + ret i32 %i0 + +else: + %i1 = call i32 @f(i32 1) + ret i32 %i1 +} + +define i32 @one_case_1(i32 %x) { +; CHECK-LABEL: @one_case_1( +; CHECK-NEXT: entry: +; CHECK-NEXT: [[SWITCH:%.*]] = icmp eq i32 [[X:%.*]], 10 +; CHECK-NEXT: br i1 [[SWITCH]], label [[A:%.*]], label [[B:%.*]] +; CHECK: common.ret: +; CHECK-NEXT: [[COMMON_RET_OP:%.*]] = phi i32 [ [[TMP0:%.*]], [[B]] ], [ [[TMP1:%.*]], [[A]] ] +; CHECK-NEXT: ret i32 [[COMMON_RET_OP]] +; CHECK: a: +; CHECK-NEXT: [[TMP0]] = call i32 @f(i32 0) +; CHECK-NEXT: br label [[COMMON_RET:%.*]] +; CHECK: b: +; CHECK-NEXT: [[TMP1]] = call i32 @f(i32 1) +; CHECK-NEXT: br label [[COMMON_RET]] +; +entry: + switch i32 %x, label %unreachable [ + i32 5, label %a + i32 6, label %a + i32 7, label %a + i32 10, label %b + ] + +unreachable: + unreachable +a: + %0 = call i32 @f(i32 0) + ret i32 %0 +b: + %1 = call i32 @f(i32 1) + ret i32 %1 +} + +define i32 @one_case_2(i32 %x) { +; CHECK-LABEL: @one_case_2( +; CHECK-NEXT: entry: +; CHECK-NEXT: [[SWITCH:%.*]] = icmp eq i32 [[X:%.*]], 5 +; CHECK-NEXT: br i1 [[SWITCH]], label [[A:%.*]], label [[B:%.*]] +; CHECK: common.ret: +; CHECK-NEXT: [[COMMON_RET_OP:%.*]] = phi i32 [ [[TMP0:%.*]], [[A]] ], [ [[TMP1:%.*]], [[B]] ] +; CHECK-NEXT: ret i32 [[COMMON_RET_OP]] +; CHECK: a: +; CHECK-NEXT: [[TMP0]] = call i32 @f(i32 0) +; CHECK-NEXT: br label [[COMMON_RET:%.*]] +; CHECK: b: +; CHECK-NEXT: [[TMP1]] = call i32 @f(i32 1) +; CHECK-NEXT: br label [[COMMON_RET]] +; +entry: + switch i32 %x, label %unreachable [ + i32 5, label %a + i32 10, label %b + i32 11, label %b + i32 12, label %b + i32 13, label %b + ] + +unreachable: + unreachable +a: + %0 = call i32 @f(i32 0) + ret i32 %0 +b: + %1 = call i32 @f(i32 1) + ret i32 %1 +} + declare void @bar(ptr nonnull dereferenceable(4)) diff --git a/llvm/test/tools/llvm-exegesis/AArch64/no-aliasing-ld-str.s b/llvm/test/tools/llvm-exegesis/AArch64/no-aliasing-ld-str.s new file mode 100644 index 0000000..65e1203 --- /dev/null +++ b/llvm/test/tools/llvm-exegesis/AArch64/no-aliasing-ld-str.s @@ -0,0 +1,8 @@ +REQUIRES: aarch64-registered-target + +RUN: llvm-exegesis -mtriple=aarch64 -mcpu=neoverse-v2 -mode=latency --dump-object-to-disk=%d --opcode-name=FMOVWSr --benchmark-phase=assemble-measured-code 2>&1 +RUN: llvm-objdump -d %d > %t.s +RUN: FileCheck %s < %t.s + +CHECK-NOT: ld{{[1-4]}} +CHECK-NOT: st{{[1-4]}} diff --git a/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2.s b/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2.s index d777d31..8e0d47e 100644 --- a/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2.s +++ b/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2.s @@ -153,12 +153,12 @@ vpshrdw $1, (%rax), %zmm17, %zmm19 {k1}{z} # CHECK-NEXT: 2 8 1.00 * vpcompressw %zmm16, (%rax) {%k1} # CHECK-NEXT: 1 1 1.00 vpcompressw %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 1 1 1.00 U vpexpandb %zmm16, %zmm19 -# CHECK-NEXT: 2 8 1.00 U vpexpandb (%rax), %zmm19 +# CHECK-NEXT: 2 8 1.00 * U vpexpandb (%rax), %zmm19 # CHECK-NEXT: 1 1 1.00 vpexpandb %zmm16, %zmm19 {%k1} # CHECK-NEXT: 2 8 1.00 * vpexpandb (%rax), %zmm19 {%k1} # CHECK-NEXT: 1 1 1.00 vpexpandb %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 1 1 1.00 U vpexpandw %zmm16, %zmm19 -# CHECK-NEXT: 2 8 1.00 U vpexpandw (%rax), %zmm19 +# CHECK-NEXT: 2 8 1.00 * U vpexpandw (%rax), %zmm19 # CHECK-NEXT: 1 1 1.00 vpexpandw %zmm16, %zmm19 {%k1} # CHECK-NEXT: 2 8 1.00 * vpexpandw (%rax), %zmm19 {%k1} # CHECK-NEXT: 1 1 1.00 vpexpandw %zmm16, %zmm19 {%k1} {z} diff --git a/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2vl.s b/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2vl.s index 99b88fe..f6be964 100644 --- a/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2vl.s +++ b/llvm/test/tools/llvm-mca/X86/Generic/resources-avx512vbmi2vl.s @@ -295,22 +295,22 @@ vpshrdw $1, (%rax), %ymm17, %ymm19 {k1}{z} # CHECK-NEXT: 2 8 1.00 * vpcompressw %ymm16, (%rax) {%k1} # CHECK-NEXT: 1 1 1.00 vpcompressw %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 1 1 1.00 U vpexpandb %xmm16, %xmm19 -# CHECK-NEXT: 2 8 1.00 U vpexpandb (%rax), %xmm19 +# CHECK-NEXT: 2 8 1.00 * U vpexpandb (%rax), %xmm19 # CHECK-NEXT: 1 1 1.00 vpexpandb %xmm16, %xmm19 {%k1} # CHECK-NEXT: 2 8 1.00 * vpexpandb (%rax), %xmm19 {%k1} # CHECK-NEXT: 1 1 1.00 vpexpandb %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 1 1 1.00 U vpexpandb %ymm16, %ymm19 -# CHECK-NEXT: 2 8 1.00 U vpexpandb (%rax), %ymm19 +# CHECK-NEXT: 2 8 1.00 * U vpexpandb (%rax), %ymm19 # CHECK-NEXT: 1 1 1.00 vpexpandb %ymm16, %ymm19 {%k1} # CHECK-NEXT: 2 8 1.00 * vpexpandb (%rax), %ymm19 {%k1} # CHECK-NEXT: 1 1 1.00 vpexpandb %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 1 1 1.00 U vpexpandw %xmm16, %xmm19 -# CHECK-NEXT: 2 8 1.00 U vpexpandw (%rax), %xmm19 +# CHECK-NEXT: 2 8 1.00 * U vpexpandw (%rax), %xmm19 # CHECK-NEXT: 1 1 1.00 vpexpandw %xmm16, %xmm19 {%k1} # CHECK-NEXT: 2 8 1.00 * vpexpandw (%rax), %xmm19 {%k1} # CHECK-NEXT: 1 1 1.00 vpexpandw %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 1 1 1.00 U vpexpandw %ymm16, %ymm19 -# CHECK-NEXT: 2 8 1.00 U vpexpandw (%rax), %ymm19 +# CHECK-NEXT: 2 8 1.00 * U vpexpandw (%rax), %ymm19 # CHECK-NEXT: 1 1 1.00 vpexpandw %ymm16, %ymm19 {%k1} # CHECK-NEXT: 2 8 1.00 * vpexpandw (%rax), %ymm19 {%k1} # CHECK-NEXT: 1 1 1.00 vpexpandw %ymm16, %ymm19 {%k1} {z} diff --git a/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2.s b/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2.s index 08f07dc..5c987ee 100644 --- a/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2.s +++ b/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2.s @@ -153,12 +153,12 @@ vpshrdw $1, (%rax), %zmm17, %zmm19 {k1}{z} # CHECK-NEXT: 2 10 1.00 * vpcompressw %zmm16, (%rax) {%k1} # CHECK-NEXT: 1 3 1.00 vpcompressw %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 1 3 1.00 U vpexpandb %zmm16, %zmm19 -# CHECK-NEXT: 2 10 1.00 U vpexpandb (%rax), %zmm19 +# CHECK-NEXT: 2 10 1.00 * U vpexpandb (%rax), %zmm19 # CHECK-NEXT: 1 3 1.00 vpexpandb %zmm16, %zmm19 {%k1} # CHECK-NEXT: 2 10 1.00 * vpexpandb (%rax), %zmm19 {%k1} # CHECK-NEXT: 1 3 1.00 vpexpandb %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 1 3 1.00 U vpexpandw %zmm16, %zmm19 -# CHECK-NEXT: 2 10 1.00 U vpexpandw (%rax), %zmm19 +# CHECK-NEXT: 2 10 1.00 * U vpexpandw (%rax), %zmm19 # CHECK-NEXT: 1 3 1.00 vpexpandw %zmm16, %zmm19 {%k1} # CHECK-NEXT: 2 10 1.00 * vpexpandw (%rax), %zmm19 {%k1} # CHECK-NEXT: 1 3 1.00 vpexpandw %zmm16, %zmm19 {%k1} {z} diff --git a/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2vl.s b/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2vl.s index 0194303..023026b 100644 --- a/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2vl.s +++ b/llvm/test/tools/llvm-mca/X86/IceLakeServer/resources-avx512vbmi2vl.s @@ -295,22 +295,22 @@ vpshrdw $1, (%rax), %ymm17, %ymm19 {k1}{z} # CHECK-NEXT: 2 10 1.00 * vpcompressw %ymm16, (%rax) {%k1} # CHECK-NEXT: 1 3 1.00 vpcompressw %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 1 3 1.00 U vpexpandb %xmm16, %xmm19 -# CHECK-NEXT: 2 10 1.00 U vpexpandb (%rax), %xmm19 +# CHECK-NEXT: 2 10 1.00 * U vpexpandb (%rax), %xmm19 # CHECK-NEXT: 1 3 1.00 vpexpandb %xmm16, %xmm19 {%k1} # CHECK-NEXT: 2 10 1.00 * vpexpandb (%rax), %xmm19 {%k1} # CHECK-NEXT: 1 3 1.00 vpexpandb %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 1 3 1.00 U vpexpandb %ymm16, %ymm19 -# CHECK-NEXT: 2 10 1.00 U vpexpandb (%rax), %ymm19 +# CHECK-NEXT: 2 10 1.00 * U vpexpandb (%rax), %ymm19 # CHECK-NEXT: 1 3 1.00 vpexpandb %ymm16, %ymm19 {%k1} # CHECK-NEXT: 2 10 1.00 * vpexpandb (%rax), %ymm19 {%k1} # CHECK-NEXT: 1 3 1.00 vpexpandb %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 1 3 1.00 U vpexpandw %xmm16, %xmm19 -# CHECK-NEXT: 2 10 1.00 U vpexpandw (%rax), %xmm19 +# CHECK-NEXT: 2 10 1.00 * U vpexpandw (%rax), %xmm19 # CHECK-NEXT: 1 3 1.00 vpexpandw %xmm16, %xmm19 {%k1} # CHECK-NEXT: 2 10 1.00 * vpexpandw (%rax), %xmm19 {%k1} # CHECK-NEXT: 1 3 1.00 vpexpandw %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 1 3 1.00 U vpexpandw %ymm16, %ymm19 -# CHECK-NEXT: 2 10 1.00 U vpexpandw (%rax), %ymm19 +# CHECK-NEXT: 2 10 1.00 * U vpexpandw (%rax), %ymm19 # CHECK-NEXT: 1 3 1.00 vpexpandw %ymm16, %ymm19 {%k1} # CHECK-NEXT: 2 10 1.00 * vpexpandw (%rax), %ymm19 {%k1} # CHECK-NEXT: 1 3 1.00 vpexpandw %ymm16, %ymm19 {%k1} {z} diff --git a/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2.s b/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2.s index ed8a417..db1f9af 100644 --- a/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2.s +++ b/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2.s @@ -153,12 +153,12 @@ vpshrdw $1, (%rax), %zmm17, %zmm19 {k1}{z} # CHECK-NEXT: 6 14 2.00 * vpcompressw %zmm16, (%rax) {%k1} # CHECK-NEXT: 2 6 2.00 vpcompressw %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 2 3 2.00 U vpexpandb %zmm16, %zmm19 -# CHECK-NEXT: 3 11 2.00 U vpexpandb (%rax), %zmm19 +# CHECK-NEXT: 3 11 2.00 * U vpexpandb (%rax), %zmm19 # CHECK-NEXT: 2 8 2.00 vpexpandb %zmm16, %zmm19 {%k1} # CHECK-NEXT: 3 13 2.00 * vpexpandb (%rax), %zmm19 {%k1} # CHECK-NEXT: 2 8 2.00 vpexpandb %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 2 3 2.00 U vpexpandw %zmm16, %zmm19 -# CHECK-NEXT: 3 11 2.00 U vpexpandw (%rax), %zmm19 +# CHECK-NEXT: 3 11 2.00 * U vpexpandw (%rax), %zmm19 # CHECK-NEXT: 2 8 2.00 vpexpandw %zmm16, %zmm19 {%k1} # CHECK-NEXT: 3 13 2.00 * vpexpandw (%rax), %zmm19 {%k1} # CHECK-NEXT: 2 8 2.00 vpexpandw %zmm16, %zmm19 {%k1} {z} diff --git a/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2vl.s b/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2vl.s index 3db09bc..9277a91 100644 --- a/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2vl.s +++ b/llvm/test/tools/llvm-mca/X86/SapphireRapids/resources-avx512vbmi2vl.s @@ -295,22 +295,22 @@ vpshrdw $1, (%rax), %ymm17, %ymm19 {k1}{z} # CHECK-NEXT: 6 14 2.00 * vpcompressw %ymm16, (%rax) {%k1} # CHECK-NEXT: 2 6 2.00 vpcompressw %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 2 3 2.00 U vpexpandb %xmm16, %xmm19 -# CHECK-NEXT: 3 10 2.00 U vpexpandb (%rax), %xmm19 +# CHECK-NEXT: 3 10 2.00 * U vpexpandb (%rax), %xmm19 # CHECK-NEXT: 2 8 2.00 vpexpandb %xmm16, %xmm19 {%k1} # CHECK-NEXT: 3 13 2.00 * vpexpandb (%rax), %xmm19 {%k1} # CHECK-NEXT: 2 8 2.00 vpexpandb %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 2 3 2.00 U vpexpandb %ymm16, %ymm19 -# CHECK-NEXT: 3 11 2.00 U vpexpandb (%rax), %ymm19 +# CHECK-NEXT: 3 11 2.00 * U vpexpandb (%rax), %ymm19 # CHECK-NEXT: 2 8 2.00 vpexpandb %ymm16, %ymm19 {%k1} # CHECK-NEXT: 3 13 2.00 * vpexpandb (%rax), %ymm19 {%k1} # CHECK-NEXT: 2 8 2.00 vpexpandb %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 2 3 2.00 U vpexpandw %xmm16, %xmm19 -# CHECK-NEXT: 3 10 2.00 U vpexpandw (%rax), %xmm19 +# CHECK-NEXT: 3 10 2.00 * U vpexpandw (%rax), %xmm19 # CHECK-NEXT: 2 8 2.00 vpexpandw %xmm16, %xmm19 {%k1} # CHECK-NEXT: 3 13 2.00 * vpexpandw (%rax), %xmm19 {%k1} # CHECK-NEXT: 2 8 2.00 vpexpandw %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 2 3 2.00 U vpexpandw %ymm16, %ymm19 -# CHECK-NEXT: 3 11 2.00 U vpexpandw (%rax), %ymm19 +# CHECK-NEXT: 3 11 2.00 * U vpexpandw (%rax), %ymm19 # CHECK-NEXT: 2 8 2.00 vpexpandw %ymm16, %ymm19 {%k1} # CHECK-NEXT: 3 13 2.00 * vpexpandw (%rax), %ymm19 {%k1} # CHECK-NEXT: 2 8 2.00 vpexpandw %ymm16, %ymm19 {%k1} {z} diff --git a/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2.s b/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2.s index 594518d..88e140d 100644 --- a/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2.s +++ b/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2.s @@ -153,12 +153,12 @@ vpshrdw $1, (%rax), %zmm17, %zmm19 {k1}{z} # CHECK-NEXT: 2 8 0.50 * vpcompressw %zmm16, (%rax) {%k1} # CHECK-NEXT: 1 5 1.00 vpcompressw %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 1 5 1.00 U vpexpandb %zmm16, %zmm19 -# CHECK-NEXT: 2 8 0.50 U vpexpandb (%rax), %zmm19 +# CHECK-NEXT: 2 8 0.50 * U vpexpandb (%rax), %zmm19 # CHECK-NEXT: 1 5 1.00 vpexpandb %zmm16, %zmm19 {%k1} # CHECK-NEXT: 2 8 0.50 * vpexpandb (%rax), %zmm19 {%k1} # CHECK-NEXT: 1 5 1.00 vpexpandb %zmm16, %zmm19 {%k1} {z} # CHECK-NEXT: 1 5 1.00 U vpexpandw %zmm16, %zmm19 -# CHECK-NEXT: 2 8 0.50 U vpexpandw (%rax), %zmm19 +# CHECK-NEXT: 2 8 0.50 * U vpexpandw (%rax), %zmm19 # CHECK-NEXT: 1 5 1.00 vpexpandw %zmm16, %zmm19 {%k1} # CHECK-NEXT: 2 8 0.50 * vpexpandw (%rax), %zmm19 {%k1} # CHECK-NEXT: 1 5 1.00 vpexpandw %zmm16, %zmm19 {%k1} {z} diff --git a/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2vl.s b/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2vl.s index 7b9c2516..325835a 100644 --- a/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2vl.s +++ b/llvm/test/tools/llvm-mca/X86/Znver4/resources-avx512vbmi2vl.s @@ -295,22 +295,22 @@ vpshrdw $1, (%rax), %ymm17, %ymm19 {k1}{z} # CHECK-NEXT: 2 8 0.50 * vpcompressw %ymm16, (%rax) {%k1} # CHECK-NEXT: 1 4 1.00 vpcompressw %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 2 1 0.50 U vpexpandb %xmm16, %xmm19 -# CHECK-NEXT: 2 8 0.50 U vpexpandb (%rax), %xmm19 +# CHECK-NEXT: 2 8 0.50 * U vpexpandb (%rax), %xmm19 # CHECK-NEXT: 2 1 0.50 vpexpandb %xmm16, %xmm19 {%k1} # CHECK-NEXT: 2 8 0.50 * vpexpandb (%rax), %xmm19 {%k1} # CHECK-NEXT: 2 1 0.50 vpexpandb %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 1 4 1.00 U vpexpandb %ymm16, %ymm19 -# CHECK-NEXT: 2 8 0.50 U vpexpandb (%rax), %ymm19 +# CHECK-NEXT: 2 8 0.50 * U vpexpandb (%rax), %ymm19 # CHECK-NEXT: 1 4 1.00 vpexpandb %ymm16, %ymm19 {%k1} # CHECK-NEXT: 2 8 0.50 * vpexpandb (%rax), %ymm19 {%k1} # CHECK-NEXT: 1 4 1.00 vpexpandb %ymm16, %ymm19 {%k1} {z} # CHECK-NEXT: 2 1 0.50 U vpexpandw %xmm16, %xmm19 -# CHECK-NEXT: 2 8 0.50 U vpexpandw (%rax), %xmm19 +# CHECK-NEXT: 2 8 0.50 * U vpexpandw (%rax), %xmm19 # CHECK-NEXT: 2 1 0.50 vpexpandw %xmm16, %xmm19 {%k1} # CHECK-NEXT: 2 8 0.50 * vpexpandw (%rax), %xmm19 {%k1} # CHECK-NEXT: 2 1 0.50 vpexpandw %xmm16, %xmm19 {%k1} {z} # CHECK-NEXT: 1 4 1.00 U vpexpandw %ymm16, %ymm19 -# CHECK-NEXT: 2 8 0.50 U vpexpandw (%rax), %ymm19 +# CHECK-NEXT: 2 8 0.50 * U vpexpandw (%rax), %ymm19 # CHECK-NEXT: 1 4 1.00 vpexpandw %ymm16, %ymm19 {%k1} # CHECK-NEXT: 2 8 0.50 * vpexpandw (%rax), %ymm19 {%k1} # CHECK-NEXT: 1 4 1.00 vpexpandw %ymm16, %ymm19 {%k1} {z} |