@@ -13,10 +13,10 @@ using namespace sycl::ext::oneapi;
1313// CHECK-NEXT: [[REF_TMP:%.*]] = alloca %"class.sycl::_V1::vec", align 8
1414// CHECK-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds nuw %"class.sycl::_V1::vec", ptr addrspace(4) [[BUF:%.*]], i64 [[ID:%.*]]
1515// CHECK-NEXT: [[TMP0:%.*]] = load i64, ptr addrspace(4) [[ARRAYIDX]], align 8
16- // CHECK-NEXT: call void @llvm.lifetime.start.p0(i64 8, ptr nonnull [[REF_TMP]]) #[[ATTR5:[0-9]+]]
16+ // CHECK-NEXT: call void @llvm.lifetime.start.p0(ptr nonnull [[REF_TMP]]) #[[ATTR5:[0-9]+]]
1717// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META8:![0-9]+]])
1818// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META11:![0-9]+]])
19- // CHECK-NEXT: call void @llvm.lifetime.start.p0(i64 8, ptr nonnull [[AGG_TMP14_I]]), !noalias [[META8]]
19+ // CHECK-NEXT: call void @llvm.lifetime.start.p0(ptr nonnull [[AGG_TMP14_I]]), !noalias [[META8]]
2020// CHECK-NEXT: store i64 [[TMP0]], ptr [[AGG_TMP14_I]], align 8, !noalias [[META8]]
2121// CHECK-NEXT: br label [[FOR_COND_I_I:%.*]]
2222// CHECK: for.cond.i.i:
@@ -33,10 +33,10 @@ using namespace sycl::ext::oneapi;
3333// CHECK-NEXT: [[INC_I_I]] = add nuw nsw i32 [[S_0_I_I]], 1
3434// CHECK-NEXT: br label [[FOR_COND_I_I]], !llvm.loop [[LOOP22:![0-9]+]]
3535// CHECK: _ZN4sycl3_V117select_from_groupINS0_9sub_groupENS0_3vecINS0_3ext6oneapi8bfloat16ELi4EEEEENSt9enable_ifIXaaoosr3stdE9is_same_vINSt5decayIT_E4typeES2_Esr4sycl3ext6oneapi12experimentalE27is_user_constructed_group_vISC_Eoosr3stdE23is_trivially_copyable_vIT0_Esr6detail6is_vecISD_EE5valueESD_E4typeESA_SD_NSA_7id_typeE.exit:
36- // CHECK-NEXT: call void @llvm.lifetime.end.p0(i64 8, ptr nonnull [[AGG_TMP14_I]]), !noalias [[META8]]
36+ // CHECK-NEXT: call void @llvm.lifetime.end.p0(ptr nonnull [[AGG_TMP14_I]]), !noalias [[META8]]
3737// CHECK-NEXT: [[TMP2:%.*]] = load i64, ptr [[REF_TMP]], align 8
3838// CHECK-NEXT: store i64 [[TMP2]], ptr addrspace(4) [[ARRAYIDX]], align 8
39- // CHECK-NEXT: call void @llvm.lifetime.end.p0(i64 8, ptr nonnull [[REF_TMP]]) #[[ATTR5]]
39+ // CHECK-NEXT: call void @llvm.lifetime.end.p0(ptr nonnull [[REF_TMP]]) #[[ATTR5]]
4040// CHECK-NEXT: ret void
4141//
4242SYCL_EXTERNAL void test_shuffle1 (sycl::sub_group &sg, vec<bfloat16, 4 > *buf,
@@ -51,10 +51,10 @@ SYCL_EXTERNAL void test_shuffle1(sycl::sub_group &sg, vec<bfloat16, 4> *buf,
5151// CHECK-NEXT: [[REF_TMP:%.*]] = alloca %"class.sycl::_V1::marray", align 2
5252// CHECK-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds nuw %"class.sycl::_V1::marray", ptr addrspace(4) [[BUF:%.*]], i64 [[ID:%.*]]
5353// CHECK-NEXT: [[TMP0:%.*]] = load i64, ptr addrspace(4) [[ARRAYIDX]], align 2, !tbaa [[TBAA25:![0-9]+]]
54- // CHECK-NEXT: call void @llvm.lifetime.start.p0(i64 8, ptr nonnull [[REF_TMP]]) #[[ATTR5]]
54+ // CHECK-NEXT: call void @llvm.lifetime.start.p0(ptr nonnull [[REF_TMP]]) #[[ATTR5]]
5555// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META26:![0-9]+]])
5656// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META29:![0-9]+]])
57- // CHECK-NEXT: call void @llvm.lifetime.start.p0(i64 8, ptr nonnull [[AGG_TMP14_I]]), !noalias [[META26]]
57+ // CHECK-NEXT: call void @llvm.lifetime.start.p0(ptr nonnull [[AGG_TMP14_I]]), !noalias [[META26]]
5858// CHECK-NEXT: store i64 [[TMP0]], ptr [[AGG_TMP14_I]], align 8, !noalias [[META26]]
5959// CHECK-NEXT: br label [[ARRAYINIT_BODY_I_I_I:%.*]]
6060// CHECK: arrayinit.body.i.i.i:
@@ -78,10 +78,10 @@ SYCL_EXTERNAL void test_shuffle1(sycl::sub_group &sg, vec<bfloat16, 4> *buf,
7878// CHECK-NEXT: [[INC_I_I]] = add nuw nsw i32 [[S_0_I_I]], 1
7979// CHECK-NEXT: br label [[FOR_COND_I_I]], !llvm.loop [[LOOP36:![0-9]+]]
8080// CHECK: _ZN4sycl3_V117select_from_groupINS0_9sub_groupENS0_6marrayINS0_3ext6oneapi8bfloat16ELm4EEEEENSt9enable_ifIXaaoosr3stdE9is_same_vINSt5decayIT_E4typeES2_Esr4sycl3ext6oneapi12experimentalE27is_user_constructed_group_vISC_Eoosr3stdE23is_trivially_copyable_vIT0_Esr6detail6is_vecISD_EE5valueESD_E4typeESA_SD_NSA_7id_typeE.exit:
81- // CHECK-NEXT: call void @llvm.lifetime.end.p0(i64 8, ptr nonnull [[AGG_TMP14_I]]), !noalias [[META26]]
81+ // CHECK-NEXT: call void @llvm.lifetime.end.p0(ptr nonnull [[AGG_TMP14_I]]), !noalias [[META26]]
8282// CHECK-NEXT: [[TMP2:%.*]] = load i64, ptr [[REF_TMP]], align 2, !tbaa [[TBAA25]]
8383// CHECK-NEXT: store i64 [[TMP2]], ptr addrspace(4) [[ARRAYIDX]], align 2, !tbaa [[TBAA25]]
84- // CHECK-NEXT: call void @llvm.lifetime.end.p0(i64 8, ptr nonnull [[REF_TMP]]) #[[ATTR5]]
84+ // CHECK-NEXT: call void @llvm.lifetime.end.p0(ptr nonnull [[REF_TMP]]) #[[ATTR5]]
8585// CHECK-NEXT: ret void
8686//
8787SYCL_EXTERNAL void test_shuffle2 (sycl::sub_group &sg, marray<bfloat16, 4 > *buf,
@@ -95,8 +95,8 @@ SYCL_EXTERNAL void test_shuffle2(sycl::sub_group &sg, marray<bfloat16, 4> *buf,
9595// CHECK-NEXT: [[AGG_TMP14_I:%.*]] = alloca %"class.sycl::_V1::marray.32", align 8
9696// CHECK-NEXT: [[REF_TMP:%.*]] = alloca %"class.sycl::_V1::marray.32", align 2
9797// CHECK-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds nuw %"class.sycl::_V1::marray.32", ptr addrspace(4) [[BUF:%.*]], i64 [[ID:%.*]]
98- // CHECK-NEXT: call void @llvm.lifetime.start.p0(i64 10, ptr nonnull [[REF_TMP]]) #[[ATTR5]]
99- // CHECK-NEXT: call void @llvm.lifetime.start.p0(i64 10, ptr nonnull [[AGG_TMP14_I]]), !noalias [[META38:![0-9]+]]
98+ // CHECK-NEXT: call void @llvm.lifetime.start.p0(ptr nonnull [[REF_TMP]]) #[[ATTR5]]
99+ // CHECK-NEXT: call void @llvm.lifetime.start.p0(ptr nonnull [[AGG_TMP14_I]]), !noalias [[META38:![0-9]+]]
100100// CHECK-NEXT: call void @llvm.memcpy.p0.p4.i64(ptr noundef nonnull align 8 dereferenceable(10) [[AGG_TMP14_I]], ptr addrspace(4) noundef align 2 dereferenceable(10) [[ARRAYIDX]], i64 10, i1 false)
101101// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META38]])
102102// CHECK-NEXT: tail call void @llvm.experimental.noalias.scope.decl(metadata [[META41:![0-9]+]])
@@ -122,9 +122,9 @@ SYCL_EXTERNAL void test_shuffle2(sycl::sub_group &sg, marray<bfloat16, 4> *buf,
122122// CHECK-NEXT: [[INC_I_I]] = add nuw nsw i32 [[S_0_I_I]], 1
123123// CHECK-NEXT: br label [[FOR_COND_I_I]], !llvm.loop [[LOOP48:![0-9]+]]
124124// CHECK: _ZN4sycl3_V117select_from_groupINS0_9sub_groupENS0_6marrayINS0_3ext6oneapi8bfloat16ELm5EEEEENSt9enable_ifIXaaoosr3stdE9is_same_vINSt5decayIT_E4typeES2_Esr4sycl3ext6oneapi12experimentalE27is_user_constructed_group_vISC_Eoosr3stdE23is_trivially_copyable_vIT0_Esr6detail6is_vecISD_EE5valueESD_E4typeESA_SD_NSA_7id_typeE.exit:
125- // CHECK-NEXT: call void @llvm.lifetime.end.p0(i64 10, ptr nonnull [[AGG_TMP14_I]]), !noalias [[META38]]
125+ // CHECK-NEXT: call void @llvm.lifetime.end.p0(ptr nonnull [[AGG_TMP14_I]]), !noalias [[META38]]
126126// CHECK-NEXT: call void @llvm.memcpy.p4.p0.i64(ptr addrspace(4) align 2 [[ARRAYIDX]], ptr align 2 [[REF_TMP]], i64 10, i1 false), !tbaa.struct [[TBAA_STRUCT49:![0-9]+]]
127- // CHECK-NEXT: call void @llvm.lifetime.end.p0(i64 10, ptr nonnull [[REF_TMP]]) #[[ATTR5]]
127+ // CHECK-NEXT: call void @llvm.lifetime.end.p0(ptr nonnull [[REF_TMP]]) #[[ATTR5]]
128128// CHECK-NEXT: ret void
129129//
130130SYCL_EXTERNAL void test_shuffle3 (sycl::sub_group &sg, marray<bfloat16, 5 > *buf,
0 commit comments