xref: /llvm-project/clang/test/OpenMP/target_has_device_addr_codegen_01.cpp (revision 94473f4db6a6f5f12d7c4081455b5b596094eac5) 
1 // NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --function-signature --include-generated-funcs --replace-value-regex "__omp_offloading_[0-9a-z]+_[0-9a-z]+" --prefix-filecheck-ir-name _
2 // RUN: %clang_cc1 -DCK1 -verify -Wno-vla -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s
3 
4 // RUN: %clang_cc1 -DCK1 -verify -Wno-vla -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s
5 // expected-no-diagnostics
6 
7 struct S {
8   int a = 0;
9   int *ptr = &a;
10   int &ref = a;
11   int arr[4];
12   S() {}
13   void foo() {
14 #pragma omp target has_device_addr(a, ref, ptr[0:4], arr[:a])
15     ++a, ++*ptr, ++ref, ++arr[0];
16   }
17 };
18 
19 int main() {
20   float a = 0;
21   float *ptr = &a;
22   float &ref = a;
23   float arr[4];
24   float vla[(int)a];
25   S s;
26   s.foo();
27 #pragma omp target has_device_addr(a, ref, ptr[0:4], arr[:(int)a], vla[0])
28   ++a, ++*ptr, ++ref, ++arr[0], ++vla[0];
29   return a;
30 }
31 
32 
33 // CHECK-LABEL: define {{[^@]+}}@main
34 // CHECK-SAME: () #[[ATTR0:[0-9]+]] {
35 // CHECK-NEXT:  entry:
36 // CHECK-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
37 // CHECK-NEXT:    [[A:%.*]] = alloca float, align 4
38 // CHECK-NEXT:    [[PTR:%.*]] = alloca ptr, align 8
39 // CHECK-NEXT:    [[REF:%.*]] = alloca ptr, align 8
40 // CHECK-NEXT:    [[ARR:%.*]] = alloca [4 x float], align 4
41 // CHECK-NEXT:    [[SAVED_STACK:%.*]] = alloca ptr, align 8
42 // CHECK-NEXT:    [[__VLA_EXPR0:%.*]] = alloca i64, align 8
43 // CHECK-NEXT:    [[S:%.*]] = alloca [[STRUCT_S:%.*]], align 8
44 // CHECK-NEXT:    [[TMP:%.*]] = alloca ptr, align 8
45 // CHECK-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [6 x ptr], align 8
46 // CHECK-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [6 x ptr], align 8
47 // CHECK-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [6 x ptr], align 8
48 // CHECK-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
49 // CHECK-NEXT:    store i32 0, ptr [[RETVAL]], align 4
50 // CHECK-NEXT:    store float 0.000000e+00, ptr [[A]], align 4
51 // CHECK-NEXT:    store ptr [[A]], ptr [[PTR]], align 8
52 // CHECK-NEXT:    store ptr [[A]], ptr [[REF]], align 8
53 // CHECK-NEXT:    [[TMP0:%.*]] = load float, ptr [[A]], align 4
54 // CHECK-NEXT:    [[CONV:%.*]] = fptosi float [[TMP0]] to i32
55 // CHECK-NEXT:    [[TMP1:%.*]] = zext i32 [[CONV]] to i64
56 // CHECK-NEXT:    [[TMP2:%.*]] = call ptr @llvm.stacksave.p0()
57 // CHECK-NEXT:    store ptr [[TMP2]], ptr [[SAVED_STACK]], align 8
58 // CHECK-NEXT:    [[VLA:%.*]] = alloca float, i64 [[TMP1]], align 4
59 // CHECK-NEXT:    store i64 [[TMP1]], ptr [[__VLA_EXPR0]], align 8
60 // CHECK-NEXT:    call void @_ZN1SC1Ev(ptr noundef nonnull align 8 dereferenceable(40) [[S]])
61 // CHECK-NEXT:    call void @_ZN1S3fooEv(ptr noundef nonnull align 8 dereferenceable(40) [[S]])
62 // CHECK-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[REF]], align 8
63 // CHECK-NEXT:    store ptr [[TMP3]], ptr [[TMP]], align 8
64 // CHECK-NEXT:    [[TMP4:%.*]] = load ptr, ptr [[PTR]], align 8
65 // CHECK-NEXT:    [[TMP5:%.*]] = load ptr, ptr [[TMP]], align 8
66 // CHECK-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
67 // CHECK-NEXT:    store ptr [[A]], ptr [[TMP6]], align 8
68 // CHECK-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0
69 // CHECK-NEXT:    store ptr [[A]], ptr [[TMP7]], align 8
70 // CHECK-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0
71 // CHECK-NEXT:    store ptr null, ptr [[TMP8]], align 8
72 // CHECK-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
73 // CHECK-NEXT:    store ptr [[TMP4]], ptr [[TMP9]], align 8
74 // CHECK-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1
75 // CHECK-NEXT:    store ptr [[TMP4]], ptr [[TMP10]], align 8
76 // CHECK-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1
77 // CHECK-NEXT:    store ptr null, ptr [[TMP11]], align 8
78 // CHECK-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
79 // CHECK-NEXT:    store ptr [[TMP5]], ptr [[TMP12]], align 8
80 // CHECK-NEXT:    [[TMP13:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 2
81 // CHECK-NEXT:    store ptr [[TMP5]], ptr [[TMP13]], align 8
82 // CHECK-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2
83 // CHECK-NEXT:    store ptr null, ptr [[TMP14]], align 8
84 // CHECK-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3
85 // CHECK-NEXT:    store ptr [[ARR]], ptr [[TMP15]], align 8
86 // CHECK-NEXT:    [[TMP16:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 3
87 // CHECK-NEXT:    store ptr [[ARR]], ptr [[TMP16]], align 8
88 // CHECK-NEXT:    [[TMP17:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 3
89 // CHECK-NEXT:    store ptr null, ptr [[TMP17]], align 8
90 // CHECK-NEXT:    [[TMP18:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 4
91 // CHECK-NEXT:    store i64 [[TMP1]], ptr [[TMP18]], align 8
92 // CHECK-NEXT:    [[TMP19:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 4
93 // CHECK-NEXT:    store i64 [[TMP1]], ptr [[TMP19]], align 8
94 // CHECK-NEXT:    [[TMP20:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 4
95 // CHECK-NEXT:    store ptr null, ptr [[TMP20]], align 8
96 // CHECK-NEXT:    [[TMP21:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 5
97 // CHECK-NEXT:    store ptr [[VLA]], ptr [[TMP21]], align 8
98 // CHECK-NEXT:    [[TMP22:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 5
99 // CHECK-NEXT:    store ptr [[VLA]], ptr [[TMP22]], align 8
100 // CHECK-NEXT:    [[TMP23:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 5
101 // CHECK-NEXT:    store ptr null, ptr [[TMP23]], align 8
102 // CHECK-NEXT:    [[TMP24:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
103 // CHECK-NEXT:    [[TMP25:%.*]] = getelementptr inbounds [6 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0
104 // CHECK-NEXT:    [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0
105 // CHECK-NEXT:    store i32 3, ptr [[TMP26]], align 4
106 // CHECK-NEXT:    [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1
107 // CHECK-NEXT:    store i32 6, ptr [[TMP27]], align 4
108 // CHECK-NEXT:    [[TMP28:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2
109 // CHECK-NEXT:    store ptr [[TMP24]], ptr [[TMP28]], align 8
110 // CHECK-NEXT:    [[TMP29:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3
111 // CHECK-NEXT:    store ptr [[TMP25]], ptr [[TMP29]], align 8
112 // CHECK-NEXT:    [[TMP30:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4
113 // CHECK-NEXT:    store ptr @.offload_sizes, ptr [[TMP30]], align 8
114 // CHECK-NEXT:    [[TMP31:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5
115 // CHECK-NEXT:    store ptr @.offload_maptypes, ptr [[TMP31]], align 8
116 // CHECK-NEXT:    [[TMP32:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6
117 // CHECK-NEXT:    store ptr null, ptr [[TMP32]], align 8
118 // CHECK-NEXT:    [[TMP33:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7
119 // CHECK-NEXT:    store ptr null, ptr [[TMP33]], align 8
120 // CHECK-NEXT:    [[TMP34:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8
121 // CHECK-NEXT:    store i64 0, ptr [[TMP34]], align 8
122 // CHECK-NEXT:    [[TMP35:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9
123 // CHECK-NEXT:    store i64 0, ptr [[TMP35]], align 8
124 // CHECK-NEXT:    [[TMP36:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10
125 // CHECK-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP36]], align 4
126 // CHECK-NEXT:    [[TMP37:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11
127 // CHECK-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP37]], align 4
128 // CHECK-NEXT:    [[TMP38:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12
129 // CHECK-NEXT:    store i32 0, ptr [[TMP38]], align 4
130 // CHECK-NEXT:    [[TMP39:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1:[0-9]+]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l27.region_id, ptr [[KERNEL_ARGS]])
131 // CHECK-NEXT:    [[TMP40:%.*]] = icmp ne i32 [[TMP39]], 0
132 // CHECK-NEXT:    br i1 [[TMP40]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
133 // CHECK:       omp_offload.failed:
134 // CHECK-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l27(ptr [[A]], ptr [[TMP4]], ptr [[TMP5]], ptr [[ARR]], i64 [[TMP1]], ptr [[VLA]]) #[[ATTR4:[0-9]+]]
135 // CHECK-NEXT:    br label [[OMP_OFFLOAD_CONT]]
136 // CHECK:       omp_offload.cont:
137 // CHECK-NEXT:    [[TMP41:%.*]] = load float, ptr [[A]], align 4
138 // CHECK-NEXT:    [[CONV1:%.*]] = fptosi float [[TMP41]] to i32
139 // CHECK-NEXT:    store i32 [[CONV1]], ptr [[RETVAL]], align 4
140 // CHECK-NEXT:    [[TMP42:%.*]] = load ptr, ptr [[SAVED_STACK]], align 8
141 // CHECK-NEXT:    call void @llvm.stackrestore.p0(ptr [[TMP42]])
142 // CHECK-NEXT:    [[TMP43:%.*]] = load i32, ptr [[RETVAL]], align 4
143 // CHECK-NEXT:    ret i32 [[TMP43]]
144 //
145 //
146 // CHECK-LABEL: define {{[^@]+}}@_ZN1SC1Ev
147 // CHECK-SAME: (ptr noundef nonnull align 8 dereferenceable(40) [[THIS:%.*]]) unnamed_addr #[[ATTR2:[0-9]+]] comdat {
148 // CHECK-NEXT:  entry:
149 // CHECK-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
150 // CHECK-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
151 // CHECK-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
152 // CHECK-NEXT:    call void @_ZN1SC2Ev(ptr noundef nonnull align 8 dereferenceable(40) [[THIS1]])
153 // CHECK-NEXT:    ret void
154 //
155 //
156 // CHECK-LABEL: define {{[^@]+}}@_ZN1S3fooEv
157 // CHECK-SAME: (ptr noundef nonnull align 8 dereferenceable(40) [[THIS:%.*]]) #[[ATTR2]] comdat {
158 // CHECK-NEXT:  entry:
159 // CHECK-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
160 // CHECK-NEXT:    [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [5 x ptr], align 8
161 // CHECK-NEXT:    [[DOTOFFLOAD_PTRS:%.*]] = alloca [5 x ptr], align 8
162 // CHECK-NEXT:    [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [5 x ptr], align 8
163 // CHECK-NEXT:    [[DOTOFFLOAD_SIZES:%.*]] = alloca [5 x i64], align 8
164 // CHECK-NEXT:    [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
165 // CHECK-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
166 // CHECK-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
167 // CHECK-NEXT:    [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
168 // CHECK-NEXT:    [[REF:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 2
169 // CHECK-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[REF]], align 8
170 // CHECK-NEXT:    [[PTR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 1
171 // CHECK-NEXT:    [[ARR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 3
172 // CHECK-NEXT:    [[TMP1:%.*]] = getelementptr [4 x i32], ptr [[ARR]], i32 1
173 // CHECK-NEXT:    [[TMP2:%.*]] = ptrtoint ptr [[TMP1]] to i64
174 // CHECK-NEXT:    [[TMP3:%.*]] = ptrtoint ptr [[A]] to i64
175 // CHECK-NEXT:    [[TMP4:%.*]] = sub i64 [[TMP2]], [[TMP3]]
176 // CHECK-NEXT:    [[TMP5:%.*]] = sdiv exact i64 [[TMP4]], ptrtoint (ptr getelementptr (i8, ptr null, i32 1) to i64)
177 // CHECK-NEXT:    call void @llvm.memcpy.p0.p0.i64(ptr align 8 [[DOTOFFLOAD_SIZES]], ptr align 8 @.offload_sizes.1, i64 40, i1 false)
178 // CHECK-NEXT:    [[TMP6:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
179 // CHECK-NEXT:    store ptr [[THIS1]], ptr [[TMP6]], align 8
180 // CHECK-NEXT:    [[TMP7:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0
181 // CHECK-NEXT:    store ptr [[A]], ptr [[TMP7]], align 8
182 // CHECK-NEXT:    [[TMP8:%.*]] = getelementptr inbounds [5 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 0
183 // CHECK-NEXT:    store i64 [[TMP5]], ptr [[TMP8]], align 8
184 // CHECK-NEXT:    [[TMP9:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0
185 // CHECK-NEXT:    store ptr null, ptr [[TMP9]], align 8
186 // CHECK-NEXT:    [[TMP10:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
187 // CHECK-NEXT:    store ptr [[THIS1]], ptr [[TMP10]], align 8
188 // CHECK-NEXT:    [[TMP11:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 1
189 // CHECK-NEXT:    store ptr [[A]], ptr [[TMP11]], align 8
190 // CHECK-NEXT:    [[TMP12:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1
191 // CHECK-NEXT:    store ptr null, ptr [[TMP12]], align 8
192 // CHECK-NEXT:    [[TMP13:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
193 // CHECK-NEXT:    store ptr [[THIS1]], ptr [[TMP13]], align 8
194 // CHECK-NEXT:    [[TMP14:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 2
195 // CHECK-NEXT:    store ptr [[TMP0]], ptr [[TMP14]], align 8
196 // CHECK-NEXT:    [[TMP15:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2
197 // CHECK-NEXT:    store ptr null, ptr [[TMP15]], align 8
198 // CHECK-NEXT:    [[TMP16:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3
199 // CHECK-NEXT:    store ptr [[THIS1]], ptr [[TMP16]], align 8
200 // CHECK-NEXT:    [[TMP17:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 3
201 // CHECK-NEXT:    store ptr [[PTR]], ptr [[TMP17]], align 8
202 // CHECK-NEXT:    [[TMP18:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 3
203 // CHECK-NEXT:    store ptr null, ptr [[TMP18]], align 8
204 // CHECK-NEXT:    [[TMP19:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 4
205 // CHECK-NEXT:    store ptr [[THIS1]], ptr [[TMP19]], align 8
206 // CHECK-NEXT:    [[TMP20:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 4
207 // CHECK-NEXT:    store ptr [[ARR]], ptr [[TMP20]], align 8
208 // CHECK-NEXT:    [[TMP21:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_MAPPERS]], i64 0, i64 4
209 // CHECK-NEXT:    store ptr null, ptr [[TMP21]], align 8
210 // CHECK-NEXT:    [[TMP22:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
211 // CHECK-NEXT:    [[TMP23:%.*]] = getelementptr inbounds [5 x ptr], ptr [[DOTOFFLOAD_PTRS]], i32 0, i32 0
212 // CHECK-NEXT:    [[TMP24:%.*]] = getelementptr inbounds [5 x i64], ptr [[DOTOFFLOAD_SIZES]], i32 0, i32 0
213 // CHECK-NEXT:    [[TMP25:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 0
214 // CHECK-NEXT:    store i32 3, ptr [[TMP25]], align 4
215 // CHECK-NEXT:    [[TMP26:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 1
216 // CHECK-NEXT:    store i32 5, ptr [[TMP26]], align 4
217 // CHECK-NEXT:    [[TMP27:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 2
218 // CHECK-NEXT:    store ptr [[TMP22]], ptr [[TMP27]], align 8
219 // CHECK-NEXT:    [[TMP28:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 3
220 // CHECK-NEXT:    store ptr [[TMP23]], ptr [[TMP28]], align 8
221 // CHECK-NEXT:    [[TMP29:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 4
222 // CHECK-NEXT:    store ptr [[TMP24]], ptr [[TMP29]], align 8
223 // CHECK-NEXT:    [[TMP30:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 5
224 // CHECK-NEXT:    store ptr @.offload_maptypes.2, ptr [[TMP30]], align 8
225 // CHECK-NEXT:    [[TMP31:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 6
226 // CHECK-NEXT:    store ptr null, ptr [[TMP31]], align 8
227 // CHECK-NEXT:    [[TMP32:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 7
228 // CHECK-NEXT:    store ptr null, ptr [[TMP32]], align 8
229 // CHECK-NEXT:    [[TMP33:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 8
230 // CHECK-NEXT:    store i64 0, ptr [[TMP33]], align 8
231 // CHECK-NEXT:    [[TMP34:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 9
232 // CHECK-NEXT:    store i64 0, ptr [[TMP34]], align 8
233 // CHECK-NEXT:    [[TMP35:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 10
234 // CHECK-NEXT:    store [3 x i32] [i32 -1, i32 0, i32 0], ptr [[TMP35]], align 4
235 // CHECK-NEXT:    [[TMP36:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 11
236 // CHECK-NEXT:    store [3 x i32] zeroinitializer, ptr [[TMP36]], align 4
237 // CHECK-NEXT:    [[TMP37:%.*]] = getelementptr inbounds nuw [[STRUCT___TGT_KERNEL_ARGUMENTS]], ptr [[KERNEL_ARGS]], i32 0, i32 12
238 // CHECK-NEXT:    store i32 0, ptr [[TMP37]], align 4
239 // CHECK-NEXT:    [[TMP38:%.*]] = call i32 @__tgt_target_kernel(ptr @[[GLOB1]], i64 -1, i32 -1, i32 0, ptr @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN1S3fooEv_l14.region_id, ptr [[KERNEL_ARGS]])
240 // CHECK-NEXT:    [[TMP39:%.*]] = icmp ne i32 [[TMP38]], 0
241 // CHECK-NEXT:    br i1 [[TMP39]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
242 // CHECK:       omp_offload.failed:
243 // CHECK-NEXT:    call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN1S3fooEv_l14(ptr [[THIS1]]) #[[ATTR4]]
244 // CHECK-NEXT:    br label [[OMP_OFFLOAD_CONT]]
245 // CHECK:       omp_offload.cont:
246 // CHECK-NEXT:    ret void
247 //
248 //
249 // CHECK-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}_main_l27
250 // CHECK-SAME: (ptr noundef nonnull align 4 dereferenceable(4) [[A:%.*]], ptr noundef [[PTR:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[REF:%.*]], ptr noundef nonnull align 4 dereferenceable(16) [[ARR:%.*]], i64 noundef [[VLA:%.*]], ptr noundef nonnull align 4 dereferenceable(4) [[VLA1:%.*]]) #[[ATTR3:[0-9]+]] {
251 // CHECK-NEXT:  entry:
252 // CHECK-NEXT:    [[A_ADDR:%.*]] = alloca ptr, align 8
253 // CHECK-NEXT:    [[PTR_ADDR:%.*]] = alloca ptr, align 8
254 // CHECK-NEXT:    [[REF_ADDR:%.*]] = alloca ptr, align 8
255 // CHECK-NEXT:    [[ARR_ADDR:%.*]] = alloca ptr, align 8
256 // CHECK-NEXT:    [[VLA_ADDR:%.*]] = alloca i64, align 8
257 // CHECK-NEXT:    [[VLA_ADDR2:%.*]] = alloca ptr, align 8
258 // CHECK-NEXT:    [[TMP:%.*]] = alloca ptr, align 8
259 // CHECK-NEXT:    store ptr [[A]], ptr [[A_ADDR]], align 8
260 // CHECK-NEXT:    store ptr [[PTR]], ptr [[PTR_ADDR]], align 8
261 // CHECK-NEXT:    store ptr [[REF]], ptr [[REF_ADDR]], align 8
262 // CHECK-NEXT:    store ptr [[ARR]], ptr [[ARR_ADDR]], align 8
263 // CHECK-NEXT:    store i64 [[VLA]], ptr [[VLA_ADDR]], align 8
264 // CHECK-NEXT:    store ptr [[VLA1]], ptr [[VLA_ADDR2]], align 8
265 // CHECK-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[A_ADDR]], align 8
266 // CHECK-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[REF_ADDR]], align 8
267 // CHECK-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[ARR_ADDR]], align 8
268 // CHECK-NEXT:    [[TMP3:%.*]] = load i64, ptr [[VLA_ADDR]], align 8
269 // CHECK-NEXT:    [[TMP4:%.*]] = load ptr, ptr [[VLA_ADDR2]], align 8
270 // CHECK-NEXT:    store ptr [[TMP1]], ptr [[TMP]], align 8
271 // CHECK-NEXT:    [[TMP5:%.*]] = load float, ptr [[TMP0]], align 4
272 // CHECK-NEXT:    [[INC:%.*]] = fadd float [[TMP5]], 1.000000e+00
273 // CHECK-NEXT:    store float [[INC]], ptr [[TMP0]], align 4
274 // CHECK-NEXT:    [[TMP6:%.*]] = load ptr, ptr [[PTR_ADDR]], align 8
275 // CHECK-NEXT:    [[TMP7:%.*]] = load float, ptr [[TMP6]], align 4
276 // CHECK-NEXT:    [[INC3:%.*]] = fadd float [[TMP7]], 1.000000e+00
277 // CHECK-NEXT:    store float [[INC3]], ptr [[TMP6]], align 4
278 // CHECK-NEXT:    [[TMP8:%.*]] = load ptr, ptr [[TMP]], align 8
279 // CHECK-NEXT:    [[TMP9:%.*]] = load float, ptr [[TMP8]], align 4
280 // CHECK-NEXT:    [[INC4:%.*]] = fadd float [[TMP9]], 1.000000e+00
281 // CHECK-NEXT:    store float [[INC4]], ptr [[TMP8]], align 4
282 // CHECK-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [4 x float], ptr [[TMP2]], i64 0, i64 0
283 // CHECK-NEXT:    [[TMP10:%.*]] = load float, ptr [[ARRAYIDX]], align 4
284 // CHECK-NEXT:    [[INC5:%.*]] = fadd float [[TMP10]], 1.000000e+00
285 // CHECK-NEXT:    store float [[INC5]], ptr [[ARRAYIDX]], align 4
286 // CHECK-NEXT:    [[ARRAYIDX6:%.*]] = getelementptr inbounds float, ptr [[TMP4]], i64 0
287 // CHECK-NEXT:    [[TMP11:%.*]] = load float, ptr [[ARRAYIDX6]], align 4
288 // CHECK-NEXT:    [[INC7:%.*]] = fadd float [[TMP11]], 1.000000e+00
289 // CHECK-NEXT:    store float [[INC7]], ptr [[ARRAYIDX6]], align 4
290 // CHECK-NEXT:    ret void
291 //
292 //
293 // CHECK-LABEL: define {{[^@]+}}@_ZN1SC2Ev
294 // CHECK-SAME: (ptr noundef nonnull align 8 dereferenceable(40) [[THIS:%.*]]) unnamed_addr #[[ATTR2]] comdat {
295 // CHECK-NEXT:  entry:
296 // CHECK-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
297 // CHECK-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
298 // CHECK-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
299 // CHECK-NEXT:    [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
300 // CHECK-NEXT:    store i32 0, ptr [[A]], align 8
301 // CHECK-NEXT:    [[PTR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 1
302 // CHECK-NEXT:    [[A2:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 0
303 // CHECK-NEXT:    store ptr [[A2]], ptr [[PTR]], align 8
304 // CHECK-NEXT:    [[REF:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 2
305 // CHECK-NEXT:    [[A3:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 0
306 // CHECK-NEXT:    store ptr [[A3]], ptr [[REF]], align 8
307 // CHECK-NEXT:    ret void
308 //
309 //
310 // CHECK-LABEL: define {{[^@]+}}@{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN1S3fooEv_l14
311 // CHECK-SAME: (ptr noundef [[THIS:%.*]]) #[[ATTR3]] {
312 // CHECK-NEXT:  entry:
313 // CHECK-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
314 // CHECK-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
315 // CHECK-NEXT:    [[TMP0:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
316 // CHECK-NEXT:    [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[TMP0]], i32 0, i32 0
317 // CHECK-NEXT:    [[TMP1:%.*]] = load i32, ptr [[A]], align 8
318 // CHECK-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP1]], 1
319 // CHECK-NEXT:    store i32 [[INC]], ptr [[A]], align 8
320 // CHECK-NEXT:    [[PTR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[TMP0]], i32 0, i32 1
321 // CHECK-NEXT:    [[TMP2:%.*]] = load ptr, ptr [[PTR]], align 8
322 // CHECK-NEXT:    [[TMP3:%.*]] = load i32, ptr [[TMP2]], align 4
323 // CHECK-NEXT:    [[INC1:%.*]] = add nsw i32 [[TMP3]], 1
324 // CHECK-NEXT:    store i32 [[INC1]], ptr [[TMP2]], align 4
325 // CHECK-NEXT:    [[REF:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[TMP0]], i32 0, i32 2
326 // CHECK-NEXT:    [[TMP4:%.*]] = load ptr, ptr [[REF]], align 8
327 // CHECK-NEXT:    [[TMP5:%.*]] = load i32, ptr [[TMP4]], align 4
328 // CHECK-NEXT:    [[INC2:%.*]] = add nsw i32 [[TMP5]], 1
329 // CHECK-NEXT:    store i32 [[INC2]], ptr [[TMP4]], align 4
330 // CHECK-NEXT:    [[ARR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[TMP0]], i32 0, i32 3
331 // CHECK-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [4 x i32], ptr [[ARR]], i64 0, i64 0
332 // CHECK-NEXT:    [[TMP6:%.*]] = load i32, ptr [[ARRAYIDX]], align 8
333 // CHECK-NEXT:    [[INC3:%.*]] = add nsw i32 [[TMP6]], 1
334 // CHECK-NEXT:    store i32 [[INC3]], ptr [[ARRAYIDX]], align 8
335 // CHECK-NEXT:    ret void
336 //
337 //
338 // SIMD-ONLY0-LABEL: define {{[^@]+}}@main
339 // SIMD-ONLY0-SAME: () #[[ATTR0:[0-9]+]] {
340 // SIMD-ONLY0-NEXT:  entry:
341 // SIMD-ONLY0-NEXT:    [[RETVAL:%.*]] = alloca i32, align 4
342 // SIMD-ONLY0-NEXT:    [[A:%.*]] = alloca float, align 4
343 // SIMD-ONLY0-NEXT:    [[PTR:%.*]] = alloca ptr, align 8
344 // SIMD-ONLY0-NEXT:    [[REF:%.*]] = alloca ptr, align 8
345 // SIMD-ONLY0-NEXT:    [[ARR:%.*]] = alloca [4 x float], align 4
346 // SIMD-ONLY0-NEXT:    [[SAVED_STACK:%.*]] = alloca ptr, align 8
347 // SIMD-ONLY0-NEXT:    [[__VLA_EXPR0:%.*]] = alloca i64, align 8
348 // SIMD-ONLY0-NEXT:    [[S:%.*]] = alloca [[STRUCT_S:%.*]], align 8
349 // SIMD-ONLY0-NEXT:    [[TMP:%.*]] = alloca ptr, align 8
350 // SIMD-ONLY0-NEXT:    store i32 0, ptr [[RETVAL]], align 4
351 // SIMD-ONLY0-NEXT:    store float 0.000000e+00, ptr [[A]], align 4
352 // SIMD-ONLY0-NEXT:    store ptr [[A]], ptr [[PTR]], align 8
353 // SIMD-ONLY0-NEXT:    store ptr [[A]], ptr [[REF]], align 8
354 // SIMD-ONLY0-NEXT:    [[TMP0:%.*]] = load float, ptr [[A]], align 4
355 // SIMD-ONLY0-NEXT:    [[CONV:%.*]] = fptosi float [[TMP0]] to i32
356 // SIMD-ONLY0-NEXT:    [[TMP1:%.*]] = zext i32 [[CONV]] to i64
357 // SIMD-ONLY0-NEXT:    [[TMP2:%.*]] = call ptr @llvm.stacksave.p0()
358 // SIMD-ONLY0-NEXT:    store ptr [[TMP2]], ptr [[SAVED_STACK]], align 8
359 // SIMD-ONLY0-NEXT:    [[VLA:%.*]] = alloca float, i64 [[TMP1]], align 4
360 // SIMD-ONLY0-NEXT:    store i64 [[TMP1]], ptr [[__VLA_EXPR0]], align 8
361 // SIMD-ONLY0-NEXT:    call void @_ZN1SC1Ev(ptr noundef nonnull align 8 dereferenceable(40) [[S]])
362 // SIMD-ONLY0-NEXT:    call void @_ZN1S3fooEv(ptr noundef nonnull align 8 dereferenceable(40) [[S]])
363 // SIMD-ONLY0-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[REF]], align 8
364 // SIMD-ONLY0-NEXT:    store ptr [[TMP3]], ptr [[TMP]], align 8
365 // SIMD-ONLY0-NEXT:    [[TMP4:%.*]] = load ptr, ptr [[REF]], align 8
366 // SIMD-ONLY0-NEXT:    [[TMP5:%.*]] = load float, ptr [[A]], align 4
367 // SIMD-ONLY0-NEXT:    [[INC:%.*]] = fadd float [[TMP5]], 1.000000e+00
368 // SIMD-ONLY0-NEXT:    store float [[INC]], ptr [[A]], align 4
369 // SIMD-ONLY0-NEXT:    [[TMP6:%.*]] = load ptr, ptr [[PTR]], align 8
370 // SIMD-ONLY0-NEXT:    [[TMP7:%.*]] = load float, ptr [[TMP6]], align 4
371 // SIMD-ONLY0-NEXT:    [[INC1:%.*]] = fadd float [[TMP7]], 1.000000e+00
372 // SIMD-ONLY0-NEXT:    store float [[INC1]], ptr [[TMP6]], align 4
373 // SIMD-ONLY0-NEXT:    [[TMP8:%.*]] = load ptr, ptr [[TMP]], align 8
374 // SIMD-ONLY0-NEXT:    [[TMP9:%.*]] = load float, ptr [[TMP8]], align 4
375 // SIMD-ONLY0-NEXT:    [[INC2:%.*]] = fadd float [[TMP9]], 1.000000e+00
376 // SIMD-ONLY0-NEXT:    store float [[INC2]], ptr [[TMP8]], align 4
377 // SIMD-ONLY0-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [4 x float], ptr [[ARR]], i64 0, i64 0
378 // SIMD-ONLY0-NEXT:    [[TMP10:%.*]] = load float, ptr [[ARRAYIDX]], align 4
379 // SIMD-ONLY0-NEXT:    [[INC3:%.*]] = fadd float [[TMP10]], 1.000000e+00
380 // SIMD-ONLY0-NEXT:    store float [[INC3]], ptr [[ARRAYIDX]], align 4
381 // SIMD-ONLY0-NEXT:    [[ARRAYIDX4:%.*]] = getelementptr inbounds float, ptr [[VLA]], i64 0
382 // SIMD-ONLY0-NEXT:    [[TMP11:%.*]] = load float, ptr [[ARRAYIDX4]], align 4
383 // SIMD-ONLY0-NEXT:    [[INC5:%.*]] = fadd float [[TMP11]], 1.000000e+00
384 // SIMD-ONLY0-NEXT:    store float [[INC5]], ptr [[ARRAYIDX4]], align 4
385 // SIMD-ONLY0-NEXT:    [[TMP12:%.*]] = load float, ptr [[A]], align 4
386 // SIMD-ONLY0-NEXT:    [[CONV6:%.*]] = fptosi float [[TMP12]] to i32
387 // SIMD-ONLY0-NEXT:    store i32 [[CONV6]], ptr [[RETVAL]], align 4
388 // SIMD-ONLY0-NEXT:    [[TMP13:%.*]] = load ptr, ptr [[SAVED_STACK]], align 8
389 // SIMD-ONLY0-NEXT:    call void @llvm.stackrestore.p0(ptr [[TMP13]])
390 // SIMD-ONLY0-NEXT:    [[TMP14:%.*]] = load i32, ptr [[RETVAL]], align 4
391 // SIMD-ONLY0-NEXT:    ret i32 [[TMP14]]
392 //
393 //
394 // SIMD-ONLY0-LABEL: define {{[^@]+}}@_ZN1SC1Ev
395 // SIMD-ONLY0-SAME: (ptr noundef nonnull align 8 dereferenceable(40) [[THIS:%.*]]) unnamed_addr #[[ATTR2:[0-9]+]] comdat {
396 // SIMD-ONLY0-NEXT:  entry:
397 // SIMD-ONLY0-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
398 // SIMD-ONLY0-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
399 // SIMD-ONLY0-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
400 // SIMD-ONLY0-NEXT:    call void @_ZN1SC2Ev(ptr noundef nonnull align 8 dereferenceable(40) [[THIS1]])
401 // SIMD-ONLY0-NEXT:    ret void
402 //
403 //
404 // SIMD-ONLY0-LABEL: define {{[^@]+}}@_ZN1S3fooEv
405 // SIMD-ONLY0-SAME: (ptr noundef nonnull align 8 dereferenceable(40) [[THIS:%.*]]) #[[ATTR2]] comdat {
406 // SIMD-ONLY0-NEXT:  entry:
407 // SIMD-ONLY0-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
408 // SIMD-ONLY0-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
409 // SIMD-ONLY0-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
410 // SIMD-ONLY0-NEXT:    [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
411 // SIMD-ONLY0-NEXT:    [[TMP0:%.*]] = load i32, ptr [[A]], align 8
412 // SIMD-ONLY0-NEXT:    [[INC:%.*]] = add nsw i32 [[TMP0]], 1
413 // SIMD-ONLY0-NEXT:    store i32 [[INC]], ptr [[A]], align 8
414 // SIMD-ONLY0-NEXT:    [[PTR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 1
415 // SIMD-ONLY0-NEXT:    [[TMP1:%.*]] = load ptr, ptr [[PTR]], align 8
416 // SIMD-ONLY0-NEXT:    [[TMP2:%.*]] = load i32, ptr [[TMP1]], align 4
417 // SIMD-ONLY0-NEXT:    [[INC2:%.*]] = add nsw i32 [[TMP2]], 1
418 // SIMD-ONLY0-NEXT:    store i32 [[INC2]], ptr [[TMP1]], align 4
419 // SIMD-ONLY0-NEXT:    [[REF:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 2
420 // SIMD-ONLY0-NEXT:    [[TMP3:%.*]] = load ptr, ptr [[REF]], align 8
421 // SIMD-ONLY0-NEXT:    [[TMP4:%.*]] = load i32, ptr [[TMP3]], align 4
422 // SIMD-ONLY0-NEXT:    [[INC3:%.*]] = add nsw i32 [[TMP4]], 1
423 // SIMD-ONLY0-NEXT:    store i32 [[INC3]], ptr [[TMP3]], align 4
424 // SIMD-ONLY0-NEXT:    [[ARR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 3
425 // SIMD-ONLY0-NEXT:    [[ARRAYIDX:%.*]] = getelementptr inbounds [4 x i32], ptr [[ARR]], i64 0, i64 0
426 // SIMD-ONLY0-NEXT:    [[TMP5:%.*]] = load i32, ptr [[ARRAYIDX]], align 8
427 // SIMD-ONLY0-NEXT:    [[INC4:%.*]] = add nsw i32 [[TMP5]], 1
428 // SIMD-ONLY0-NEXT:    store i32 [[INC4]], ptr [[ARRAYIDX]], align 8
429 // SIMD-ONLY0-NEXT:    ret void
430 //
431 //
432 // SIMD-ONLY0-LABEL: define {{[^@]+}}@_ZN1SC2Ev
433 // SIMD-ONLY0-SAME: (ptr noundef nonnull align 8 dereferenceable(40) [[THIS:%.*]]) unnamed_addr #[[ATTR2]] comdat {
434 // SIMD-ONLY0-NEXT:  entry:
435 // SIMD-ONLY0-NEXT:    [[THIS_ADDR:%.*]] = alloca ptr, align 8
436 // SIMD-ONLY0-NEXT:    store ptr [[THIS]], ptr [[THIS_ADDR]], align 8
437 // SIMD-ONLY0-NEXT:    [[THIS1:%.*]] = load ptr, ptr [[THIS_ADDR]], align 8
438 // SIMD-ONLY0-NEXT:    [[A:%.*]] = getelementptr inbounds nuw [[STRUCT_S:%.*]], ptr [[THIS1]], i32 0, i32 0
439 // SIMD-ONLY0-NEXT:    store i32 0, ptr [[A]], align 8
440 // SIMD-ONLY0-NEXT:    [[PTR:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 1
441 // SIMD-ONLY0-NEXT:    [[A2:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 0
442 // SIMD-ONLY0-NEXT:    store ptr [[A2]], ptr [[PTR]], align 8
443 // SIMD-ONLY0-NEXT:    [[REF:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 2
444 // SIMD-ONLY0-NEXT:    [[A3:%.*]] = getelementptr inbounds nuw [[STRUCT_S]], ptr [[THIS1]], i32 0, i32 0
445 // SIMD-ONLY0-NEXT:    store ptr [[A3]], ptr [[REF]], align 8
446 // SIMD-ONLY0-NEXT:    ret void
447 //
448