Skip to content

Commit cb2cc2f

Browse files
e-kudjsji
e-kud
authored and
jsji
committed
Partially revert "Add nuw in GEP for CodeGenSYCL tests"
This partially reverts commit f9d8cf0.
1 parent 3b965b9 commit cb2cc2f

File tree

4 files changed

+9
-9
lines changed

4 files changed

+9
-9
lines changed

‎clang/test/CodeGenSYCL/address-space-new.cpp

Lines changed: 1 addition & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -64,7 +64,7 @@ void test() {
6464
*aptr = 44;
6565
// CHECK: [[TMP13:%.*]] = load ptr addrspace(4), ptr addrspace(4) %aptr.ascast
6666
// CHECK: [[ARRAYDECAY2:%.*]] = getelementptr inbounds [42 x i32], ptr addrspace(4) [[ARR_ASCAST]], i64 0, i64 0
67-
// CHECK: [[ADD_PTR3:%.*]] = getelementptr inbounds nuw i32, ptr addrspace(4) [[ARRAYDECAY2]], i64 168
67+
// CHECK: [[ADD_PTR3:%.*]] = getelementptr inbounds i32, ptr addrspace(4) [[ARRAYDECAY2]], i64 168
6868
// CHECK: [[CMP4:%.*]] = icmp ult ptr addrspace(4) [[TMP13]], [[ADD_PTR3]]
6969

7070
const char *str = "Hello, world!";

‎clang/test/CodeGenSYCL/kernel-param-acc-array.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -58,7 +58,7 @@ int main() {
5858
// CHECK-NEXT: [[ARRAYINIT_ELEMENT:%.*]] = getelementptr inbounds %"class.sycl::_V1::accessor", ptr addrspace(4) [[ACC]], i64 1
5959
// CHECK-NEXT: call spir_func void @_ZN4sycl3_V18accessorIiLi1ELNS0_6access4modeE1026ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_3ext6oneapi22accessor_property_listIJEEEEC1Ev(ptr addrspace(4) noundef align 4 dereferenceable_or_null(12) [[ARRAYINIT_ELEMENT]]) #[[ATTR4]]
6060
// CHECK-NEXT: [[ACC9:%.*]] = getelementptr inbounds nuw [[CLASS_ANON]], ptr addrspace(4) [[__SYCLKERNEL_ASCAST]], i32 0, i32 0
61-
// CHECK-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds nuw [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[ACC9]], i64 0, i64 0
61+
// CHECK-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[ACC9]], i64 0, i64 0
6262
// CHECK-NEXT: [[TMP0:%.*]] = load ptr addrspace(1), ptr addrspace(4) [[_ARG_ACC_ADDR_ASCAST]], align 8
6363
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP_ASCAST]], ptr addrspace(4) align 4 [[_ARG_ACC1_ASCAST]], i64 4, i1 false)
6464
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP10_ASCAST]], ptr addrspace(4) align 4 [[_ARG_ACC2_ASCAST]], i64 4, i1 false)
@@ -68,7 +68,7 @@ int main() {
6868
// CHECK-NEXT: [[AGG_TMP11_ASCAST_ASCAST:%.*]] = addrspacecast ptr addrspace(4) [[AGG_TMP11_ASCAST]] to ptr
6969
// CHECK-NEXT: call spir_func void @{{.*}}__init{{.*}}(ptr addrspace(4) noundef align 4 dereferenceable_or_null(12) [[ARRAYIDX]], ptr addrspace(1) noundef [[TMP0]], ptr noundef byval(%"struct.sycl::_V1::range") align 4 [[AGG_TMP_ASCAST_ASCAST]], ptr noundef byval(%"struct.sycl::_V1::range") align 4 [[AGG_TMP10_ASCAST_ASCAST]], ptr noundef byval(%"struct.sycl::_V1::id") align 4 [[AGG_TMP11_ASCAST_ASCAST]]) #[[ATTR4]]
7070
// CHECK-NEXT: [[ACC12:%.*]] = getelementptr inbounds nuw [[CLASS_ANON]], ptr addrspace(4) [[__SYCLKERNEL_ASCAST]], i32 0, i32 0
71-
// CHECK-NEXT: [[ARRAYIDX13:%.*]] = getelementptr inbounds nuw [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[ACC12]], i64 0, i64 1
71+
// CHECK-NEXT: [[ARRAYIDX13:%.*]] = getelementptr inbounds [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[ACC12]], i64 0, i64 1
7272
// CHECK-NEXT: [[TMP1:%.*]] = load ptr addrspace(1), ptr addrspace(4) [[_ARG_ACC_ADDR5_ASCAST]], align 8
7373
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP14_ASCAST]], ptr addrspace(4) align 4 [[_ARG_ACC6_ASCAST]], i64 4, i1 false)
7474
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP15_ASCAST]], ptr addrspace(4) align 4 [[_ARG_ACC7_ASCAST]], i64 4, i1 false)

‎clang/test/CodeGenSYCL/kernel-param-member-acc-array.cpp

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -64,7 +64,7 @@ int main() {
6464
// CHECK-NEXT: call spir_func void @_ZN4sycl3_V18accessorIiLi1ELNS0_6access4modeE1026ELNS2_6targetE2014ELNS2_11placeholderE0ENS0_3ext6oneapi22accessor_property_listIJEEEEC1Ev(ptr addrspace(4) noundef align 4 dereferenceable_or_null(12) [[ARRAYINIT_ELEMENT]]) #[[ATTR4]]
6565
// CHECK-NEXT: [[STRUCT_ACC9:%.*]] = getelementptr inbounds nuw [[CLASS_ANON]], ptr addrspace(4) [[__SYCLKERNEL_ASCAST]], i32 0, i32 0
6666
// CHECK-NEXT: [[MEMBER_ACC10:%.*]] = getelementptr inbounds nuw [[STRUCT_STRUCT_ACC_T]], ptr addrspace(4) [[STRUCT_ACC9]], i32 0, i32 0
67-
// CHECK-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds nuw [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[MEMBER_ACC10]], i64 0, i64 0
67+
// CHECK-NEXT: [[ARRAYIDX:%.*]] = getelementptr inbounds [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[MEMBER_ACC10]], i64 0, i64 0
6868
// CHECK-NEXT: [[TMP0:%.*]] = load ptr addrspace(1), ptr addrspace(4) [[_ARG_MEMBER_ACC_ADDR_ASCAST]], align 8
6969
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP_ASCAST]], ptr addrspace(4) align 4 [[_ARG_MEMBER_ACC1_ASCAST]], i64 4, i1 false)
7070
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP11_ASCAST]], ptr addrspace(4) align 4 [[_ARG_MEMBER_ACC2_ASCAST]], i64 4, i1 false)
@@ -75,7 +75,7 @@ int main() {
7575
// CHECK-NEXT: call spir_func void @{{.*}}__init{{.*}}(ptr addrspace(4) noundef align 4 dereferenceable_or_null(12) [[ARRAYIDX]], ptr addrspace(1) noundef [[TMP0]], ptr noundef byval(%"struct.sycl::_V1::range") align 4 [[AGG_TMP_ASCAST_ASCAST]], ptr noundef byval(%"struct.sycl::_V1::range") align 4 [[AGG_TMP11_ASCAST_ASCAST]], ptr noundef byval(%"struct.sycl::_V1::id") align 4 [[AGG_TMP12_ASCAST_ASCAST]]) #[[ATTR4]]
7676
// CHECK-NEXT: [[STRUCT_ACC13:%.*]] = getelementptr inbounds nuw [[CLASS_ANON]], ptr addrspace(4) [[__SYCLKERNEL_ASCAST]], i32 0, i32 0
7777
// CHECK-NEXT: [[MEMBER_ACC14:%.*]] = getelementptr inbounds nuw [[STRUCT_STRUCT_ACC_T]], ptr addrspace(4) [[STRUCT_ACC13]], i32 0, i32 0
78-
// CHECK-NEXT: [[ARRAYIDX15:%.*]] = getelementptr inbounds nuw [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[MEMBER_ACC14]], i64 0, i64 1
78+
// CHECK-NEXT: [[ARRAYIDX15:%.*]] = getelementptr inbounds [2 x %"class.sycl::_V1::accessor"], ptr addrspace(4) [[MEMBER_ACC14]], i64 0, i64 1
7979
// CHECK-NEXT: [[TMP1:%.*]] = load ptr addrspace(1), ptr addrspace(4) [[_ARG_MEMBER_ACC_ADDR5_ASCAST]], align 8
8080
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP16_ASCAST]], ptr addrspace(4) align 4 [[_ARG_MEMBER_ACC6_ASCAST]], i64 4, i1 false)
8181
// CHECK-NEXT: call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 [[AGG_TMP17_ASCAST]], ptr addrspace(4) align 4 [[_ARG_MEMBER_ACC7_ASCAST]], i64 4, i1 false)

‎clang/test/CodeGenSYCL/kernel-param-pod-array.cpp

Lines changed: 4 additions & 4 deletions
Original file line numberDiff line numberDiff line change
@@ -62,7 +62,7 @@ int main() {
6262
// CHECK: [[ARRAYINITBODY]]:
6363
// CHECK: %[[ARRAYINDEX:.+]] = phi i64 [ 0, %{{.*}} ], [ %[[NEXTINDEX:.+]], %[[ARRAYINITBODY]] ]
6464
// CHECK: %[[TARG_ARRAY_ELEM:.+]] = getelementptr inbounds i32, ptr addrspace(4) %[[ARRAY_BEGIN]], i64 %[[ARRAYINDEX]]
65-
// CHECK: %[[SRC_ELEM:.+]] = getelementptr inbounds nuw [2 x i32], ptr addrspace(4) %[[WRAPPER_PTR]], i64 0, i64 %[[ARRAYINDEX]]
65+
// CHECK: %[[SRC_ELEM:.+]] = getelementptr inbounds [2 x i32], ptr addrspace(4) %[[WRAPPER_PTR]], i64 0, i64 %[[ARRAYINDEX]]
6666
// CHECK: %[[SRC_VAL:.+]] = load i32, ptr addrspace(4) %[[SRC_ELEM]]
6767
// CHECK: store i32 %[[SRC_VAL]], ptr addrspace(4) %[[TARG_ARRAY_ELEM]]
6868
// CHECK: %[[NEXTINDEX]] = add nuw i64 %[[ARRAYINDEX]], 1
@@ -87,7 +87,7 @@ int main() {
8787
// CHECK: [[ARRAYINITBODY]]:
8888
// CHECK: %[[ARRAYINDEX:.+]] = phi i64 [ 0, %{{.*}} ], [ %[[NEXTINDEX:.+]], %[[ARRAYINITBODY]] ]
8989
// CHECK: %[[TARG_ARRAY_ELEM:.+]] = getelementptr inbounds %struct{{.*}}.foo, ptr addrspace(4) %[[ARRAY_BEGIN]], i64 %[[ARRAYINDEX]]
90-
// CHECK: %[[SRC_ELEM:.+]] = getelementptr inbounds nuw [2 x %struct{{.*}}.foo], ptr addrspace(4) %[[WRAPPER_PTR]], i64 0, i64 %[[ARRAYINDEX]]
90+
// CHECK: %[[SRC_ELEM:.+]] = getelementptr inbounds [2 x %struct{{.*}}.foo], ptr addrspace(4) %[[WRAPPER_PTR]], i64 0, i64 %[[ARRAYINDEX]]
9191
// call void @llvm.memcpy.p4.p4.i64(ptr addrspace(4) align 4 %[[TARG_ARRAY_ELEM]], ptr addrspace(4) align %[[SRC_ELEM]], i64 24, i1 false)
9292
// CHECK: %[[NEXTINDEX]] = add nuw i64 %[[ARRAYINDEX]], 1
9393
// CHECK: %[[ISDONE:.+]] = icmp eq i64 %[[NEXTINDEX]], 2
@@ -111,15 +111,15 @@ int main() {
111111
// CHECK: [[ARRAYINITBODY]]:
112112
// CHECK: %[[ARRAYINDEX:.+]] = phi i64 [ 0, %{{.*}} ], [ %[[NEXTINDEX:.+]], %[[ARRAYINITEND:.+]] ]
113113
// CHECK: %[[TARG_OUTER_ELEM:.+]] = getelementptr inbounds [1 x i32], ptr addrspace(4) %[[ARRAY_BEGIN]], i64 %[[ARRAYINDEX]]
114-
// CHECK: %[[SRC_OUTER_ELEM:.+]] = getelementptr inbounds nuw [2 x [1 x i32]], ptr addrspace(4) %[[WRAPPER_PTR]], i64 0, i64 %[[ARRAYINDEX]]
114+
// CHECK: %[[SRC_OUTER_ELEM:.+]] = getelementptr inbounds [2 x [1 x i32]], ptr addrspace(4) %[[WRAPPER_PTR]], i64 0, i64 %[[ARRAYINDEX]]
115115
// CHECK: %[[ARRAY_BEGIN_INNER:.+]] = getelementptr inbounds [1 x i32], ptr addrspace(4) %[[TARG_OUTER_ELEM]], i64 0, i64 0
116116
// CHECK: br label %[[ARRAYINITBODY_INNER:.+]]
117117

118118
// Check Inner Loop
119119
// CHECK: [[ARRAYINITBODY_INNER]]:
120120
// CHECK: %[[ARRAYINDEX_INNER:.+]] = phi i64 [ 0, %{{.*}} ], [ %[[NEXTINDEX_INNER:.+]], %[[ARRAYINITBODY_INNER:.+]] ]
121121
// CHECK: %[[TARG_INNER_ELEM:.+]] = getelementptr inbounds i32, ptr addrspace(4) %[[ARRAY_BEGIN_INNER]], i64 %[[ARRAYINDEX_INNER]]
122-
// CHECK: %[[SRC_INNER_ELEM:.+]] = getelementptr inbounds nuw [1 x i32], ptr addrspace(4) %[[SRC_OUTER_ELEM]], i64 0, i64 %[[ARRAYINDEX_INNER]]
122+
// CHECK: %[[SRC_INNER_ELEM:.+]] = getelementptr inbounds [1 x i32], ptr addrspace(4) %[[SRC_OUTER_ELEM]], i64 0, i64 %[[ARRAYINDEX_INNER]]
123123
// CHECK: %[[SRC_LOAD:.+]] = load i32, ptr addrspace(4) %[[SRC_INNER_ELEM]]
124124
// CHECK: store i32 %[[SRC_LOAD]], ptr addrspace(4) %[[TARG_INNER_ELEM]]
125125
// CHECK: %[[NEXTINDEX_INNER]] = add nuw i64 %[[ARRAYINDEX_INNER]], 1

0 commit comments

Comments
 (0)