1 // Test host codegen.
2 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64
3 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s
4 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64
5 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32
6 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s
7 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32
8 
9 // Test target codegen - host bc file has to be created first.
10 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
11 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-64
12 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s
13 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-64
14 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc
15 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-32
16 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s
17 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix TCHECK --check-prefix TCHECK-32
18 
19 // expected-no-diagnostics
20 #ifndef HEADER
21 #define HEADER
22 
23 // CHECK-DAG: %ident_t = type { i32, i32, i32, i32, i8* }
24 // CHECK-DAG: [[STR:@.+]] = private unnamed_addr constant [23 x i8] c";unknown;unknown;0;0;;\00"
25 // CHECK-DAG: [[DEF_LOC:@.+]] = private unnamed_addr constant %ident_t { i32 0, i32 2, i32 0, i32 0, i8* getelementptr inbounds ([23 x i8], [23 x i8]* [[STR]], i32 0, i32 0) }
26 
27 // CHECK-DAG: [[S1:%.+]] = type { double }
28 // CHECK-DAG: [[ENTTY:%.+]] = type { i8*, i8*, i[[SZ:32|64]], i32, i32 }
29 // CHECK-DAG: [[DEVTY:%.+]] = type { i8*, i8*, [[ENTTY]]*, [[ENTTY]]* }
30 // CHECK-DAG: [[DSCTY:%.+]] = type { i32, [[DEVTY]]*, [[ENTTY]]*, [[ENTTY]]* }
31 
32 // TCHECK: [[ENTTY:%.+]] = type { i8*, i8*, i{{32|64}}, i32, i32 }
33 
34 // CHECK-DAG: $[[REGFN:\.omp_offloading\..+]] = comdat
35 
36 // We have 6 target regions
37 
38 // CHECK-DAG: @{{.*}} = private constant i8 0
39 // CHECK-DAG: @{{.*}} = private constant i8 0
40 // CHECK-DAG: @{{.*}} = private constant i8 0
41 // CHECK-DAG: @{{.*}} = private constant i8 0
42 // CHECK-DAG: @{{.*}} = private constant i8 0
43 // CHECK-DAG: @{{.*}} = private constant i8 0
44 
45 // TCHECK: @{{.+}} = constant [[ENTTY]]
46 // TCHECK: @{{.+}} = constant [[ENTTY]]
47 // TCHECK: @{{.+}} = constant [[ENTTY]]
48 // TCHECK: @{{.+}} = constant [[ENTTY]]
49 // TCHECK: @{{.+}} = constant [[ENTTY]]
50 // TCHECK: @{{.+}} = constant [[ENTTY]]
51 
52 // Check if offloading descriptor is created.
53 // CHECK: [[ENTBEGIN:@.+]] = external constant [[ENTTY]]
54 // CHECK: [[ENTEND:@.+]] = external constant [[ENTTY]]
55 // CHECK: [[DEVBEGIN:@.+]] = external constant i8
56 // CHECK: [[DEVEND:@.+]] = external constant i8
57 // CHECK: [[IMAGES:@.+]] = internal unnamed_addr constant [1 x [[DEVTY]]] [{{.+}} { i8* [[DEVBEGIN]], i8* [[DEVEND]], [[ENTTY]]* [[ENTBEGIN]], [[ENTTY]]* [[ENTEND]] }], comdat($[[REGFN]])
58 // CHECK: [[DESC:@.+]] = internal constant [[DSCTY]] { i32 1, [[DEVTY]]* getelementptr inbounds ([1 x [[DEVTY]]], [1 x [[DEVTY]]]* [[IMAGES]], i32 0, i32 0), [[ENTTY]]* [[ENTBEGIN]], [[ENTTY]]* [[ENTEND]] }, comdat($[[REGFN]])
59 
60 // Check target registration is registered as a Ctor.
61 // CHECK: appending global [1 x { i32, void ()*, i8* }] [{ i32, void ()*, i8* } { i32 0, void ()* bitcast (void (i8*)* @[[REGFN]] to void ()*), i8* bitcast (void (i8*)* @[[REGFN]] to i8*) }]
62 
63 
64 template<typename tx>
65 tx ftemplate(int n) {
66   tx a = 0;
67 
68   #pragma omp target parallel if(parallel: 0)
69   {
70     a += 1;
71   }
72 
73   short b = 1;
74   #pragma omp target parallel if(parallel: 1)
75   {
76     a += b;
77   }
78 
79   return a;
80 }
81 
82 static
83 int fstatic(int n) {
84 
85   #pragma omp target parallel if(n>1)
86   {
87   }
88 
89   #pragma omp target parallel if(target: n-2>2)
90   {
91   }
92 
93   return n+1;
94 }
95 
96 struct S1 {
97   double a;
98 
99   int r1(int n){
100     int b = 1;
101 
102     #pragma omp target parallel if(parallel: n>3)
103     {
104       this->a = (double)b + 1.5;
105     }
106 
107     #pragma omp target parallel if(target: n>4) if(parallel: n>5)
108     {
109       this->a = 2.5;
110     }
111 
112     return (int)a;
113   }
114 };
115 
116 // CHECK: define {{.*}}@{{.*}}bar{{.*}}
117 int bar(int n){
118   int a = 0;
119 
120   S1 S;
121   // CHECK: call {{.*}}i32 [[FS1:@.+]]([[S1]]* {{.*}}, i32 {{.*}})
122   a += S.r1(n);
123 
124   // CHECK: call {{.*}}i32 [[FSTATIC:@.+]](i32 {{.*}})
125   a += fstatic(n);
126 
127   // CHECK: call {{.*}}i32 [[FTEMPLATE:@.+]](i32 {{.*}})
128   a += ftemplate<int>(n);
129 
130   return a;
131 }
132 
133 
134 
135 //
136 // CHECK: define {{.*}}[[FS1]]([[S1]]* {{%.+}}, i32 {{[^%]*}}[[PARM:%.+]])
137 //
138 // CHECK-DAG:   store i32 [[PARM]], i32* [[N_ADDR:%.+]], align
139 // CHECK:       [[NV:%.+]] = load i32, i32* [[N_ADDR]], align
140 // CHECK:       [[CMP:%.+]] = icmp sgt i32 [[NV]], 3
141 // CHECK:       [[FB:%.+]] = zext i1 [[CMP]] to i8
142 // CHECK:       store i8 [[FB]], i8* [[CAPE_ADDR:%.+]], align
143 // CHECK:       [[CAPE:%.+]] = load i8, i8* [[CAPE_ADDR]], align
144 // CHECK:       [[TB:%.+]] = trunc i8 [[CAPE]] to i1
145 // CHECK:       [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i8*
146 // CHECK:       [[FB:%.+]] = zext i1 [[TB]] to i8
147 // CHECK:       store i8 [[FB]], i8* [[CONV]], align
148 // CHECK:       [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
149 //
150 // CHECK-DAG:   [[RET:%.+]] = call i32 @__tgt_target_teams(i32 -1, i8* @{{[^,]+}}, i32 3, {{.*}}, i32 1, i32 0)
151 // CHECK:       store i32 [[RET]], i32* [[RHV:%.+]], align
152 // CHECK:       [[RET2:%.+]] = load i32, i32* [[RHV]], align
153 // CHECK:       [[ERROR:%.+]] = icmp ne i32 [[RET2]], 0
154 // CHECK:       br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
155 //
156 // CHECK:       [[FAIL]]
157 // CHECK:       call void [[HVT1:@.+]]([[S1]]* {{%.+}}, i[[SZ]] {{%.+}}, i[[SZ]] [[ARG]])
158 // CHECK:       br label {{%?}}[[END]]
159 // CHECK:       [[END]]
160 //
161 //
162 //
163 // CHECK:       [[NV:%.+]] = load i32, i32* [[N_ADDR]], align
164 // CHECK:       [[CMP:%.+]] = icmp sgt i32 [[NV]], 5
165 // CHECK:       [[FB:%.+]] = zext i1 [[CMP]] to i8
166 // CHECK:       store i8 [[FB]], i8* [[CAPE_ADDR:%.+]], align
167 // CHECK:       [[CAPE:%.+]] = load i8, i8* [[CAPE_ADDR]], align
168 // CHECK:       [[TB:%.+]] = trunc i8 [[CAPE]] to i1
169 // CHECK:       [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i8*
170 // CHECK:       [[FB:%.+]] = zext i1 [[TB]] to i8
171 // CHECK:       store i8 [[FB]], i8* [[CONV]], align
172 // CHECK:       [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
173 // CHECK:       [[NV:%.+]] = load i32, i32* [[N_ADDR]], align
174 // CHECK:       [[CMP:%.+]] = icmp sgt i32 [[NV]], 4
175 // CHECK:       br i1 [[CMP]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]]
176 //
177 // CHECK:       [[IF_THEN]]
178 // CHECK-DAG:   [[RET:%.+]] = call i32 @__tgt_target_teams(i32 -1, i8* @{{[^,]+}}, i32 2, {{.*}}, i32 1, i32 0)
179 // CHECK:       store i32 [[RET]], i32* [[RHV:%.+]], align
180 // CHECK:       br label {{%?}}[[END:.+]]
181 //
182 // CHECK:       [[IF_ELSE]]
183 // CHECK:       store i32 -1, i32* [[RHV]], align
184 // CHECK:       br label {{%?}}[[END]]
185 //
186 // CHECK:       [[END]]
187 // CHECK:       [[RET2:%.+]] = load i32, i32* [[RHV]], align
188 // CHECK:       [[ERROR:%.+]] = icmp ne i32 [[RET2]], 0
189 // CHECK:       br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
190 //
191 // CHECK:       [[FAIL]]
192 // CHECK:       call void [[HVT2:@.+]]([[S1]]* {{%.+}}, i[[SZ]] [[ARG]])
193 // CHECK:       br label {{%?}}[[END]]
194 // CHECK:       [[END]]
195 
196 
197 
198 
199 
200 
201 //
202 // CHECK: define {{.*}}[[FSTATIC]](i32 {{[^%]*}}[[PARM:%.+]])
203 //
204 // CHECK-DAG:   store i32 [[PARM]], i32* [[N_ADDR:%.+]], align
205 // CHECK:       [[NV:%.+]] = load i32, i32* [[N_ADDR]], align
206 // CHECK:       [[CMP:%.+]] = icmp sgt i32 [[NV]], 1
207 // CHECK:       [[FB:%.+]] = zext i1 [[CMP]] to i8
208 // CHECK:       store i8 [[FB]], i8* [[CAPE_ADDR:%.+]], align
209 // CHECK:       [[CAPE:%.+]] = load i8, i8* [[CAPE_ADDR]], align
210 // CHECK:       [[TB:%.+]] = trunc i8 [[CAPE]] to i1
211 // CHECK:       [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPEC_ADDR:%.+]] to i8*
212 // CHECK:       [[FB:%.+]] = zext i1 [[TB]] to i8
213 // CHECK:       store i8 [[FB]], i8* [[CONV]], align
214 // CHECK:       [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[CAPEC_ADDR]], align
215 // CHECK:       [[CAPE2:%.+]] = load i8, i8* [[CAPE_ADDR]], align
216 // CHECK:       [[TB:%.+]] = trunc i8 [[CAPE2]] to i1
217 // CHECK:       br i1 [[TB]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]]
218 //
219 // CHECK:       [[IF_THEN]]
220 // CHECK-DAG:   [[RET:%.+]] = call i32 @__tgt_target_teams(i32 -1, i8* @{{[^,]+}}, i32 1, {{.*}}, i32 1, i32 0)
221 // CHECK:       store i32 [[RET]], i32* [[RHV:%.+]], align
222 // CHECK:       br label {{%?}}[[END:.+]]
223 //
224 // CHECK:       [[IF_ELSE]]
225 // CHECK:       store i32 -1, i32* [[RHV]], align
226 // CHECK:       br label {{%?}}[[END]]
227 //
228 // CHECK:       [[END]]
229 // CHECK:       [[RET2:%.+]] = load i32, i32* [[RHV]], align
230 // CHECK:       [[ERROR:%.+]] = icmp ne i32 [[RET2]], 0
231 // CHECK:       br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
232 //
233 // CHECK:       [[FAIL]]
234 // CHECK:       call void [[HVT3:@.+]](i[[SZ]] [[ARG]])
235 // CHECK:       br label {{%?}}[[END]]
236 // CHECK:       [[END]]
237 //
238 //
239 //
240 // CHECK-DAG:   [[NV:%.+]] = load i32, i32* [[N_ADDR]], align
241 // CHECK:       [[SUB:%.+]] = sub nsw i32 [[NV]], 2
242 // CHECK:       [[CMP:%.+]] = icmp sgt i32 [[SUB]], 2
243 // CHECK:       br i1 [[CMP]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]]
244 //
245 // CHECK:       [[IF_THEN]]
246 // CHECK-DAG:   [[RET:%.+]] = call i32 @__tgt_target_teams(i32 -1, i8* @{{[^,]+}}, i32 0, {{.*}}, i32 1, i32 0)
247 // CHECK:       store i32 [[RET]], i32* [[RHV:%.+]], align
248 // CHECK:       br label {{%?}}[[END:.+]]
249 //
250 // CHECK:       [[IF_ELSE]]
251 // CHECK:       store i32 -1, i32* [[RHV]], align
252 // CHECK:       br label {{%?}}[[END]]
253 //
254 // CHECK:       [[END]]
255 // CHECK:       [[RET2:%.+]] = load i32, i32* [[RHV]], align
256 // CHECK:       [[ERROR:%.+]] = icmp ne i32 [[RET2]], 0
257 // CHECK:       br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
258 //
259 // CHECK:       [[FAIL]]
260 // CHECK:       call void [[HVT4:@.+]]()
261 // CHECK:       br label {{%?}}[[END]]
262 // CHECK:       [[END]]
263 
264 
265 
266 
267 
268 
269 //
270 // CHECK: define {{.*}}[[FTEMPLATE]]
271 //
272 // CHECK-DAG:   [[RET:%.+]] = call i32 @__tgt_target_teams(i32 -1, i8* @{{[^,]+}}, i32 1, {{.*}}, i32 1, i32 0)
273 // CHECK-NEXT:  store i32 [[RET]], i32* [[RHV:%.+]], align
274 // CHECK-NEXT:  [[RET2:%.+]] = load i32, i32* [[RHV]], align
275 // CHECK-NEXT:  [[ERROR:%.+]] = icmp ne i32 [[RET2]], 0
276 // CHECK-NEXT:  br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
277 //
278 // CHECK:       [[FAIL]]
279 // CHECK:       call void [[HVT5:@.+]]({{[^,]+}})
280 // CHECK:       br label {{%?}}[[END]]
281 //
282 // CHECK:       [[END]]
283 //
284 //
285 //
286 // CHECK-DAG:   [[RET:%.+]] = call i32 @__tgt_target_teams(i32 -1, i8* @{{[^,]+}}, i32 2, {{.*}}, i32 1, i32 0)
287 // CHECK-NEXT:  store i32 [[RET]], i32* [[RHV:%.+]], align
288 // CHECK-NEXT:  [[RET2:%.+]] = load i32, i32* [[RHV]], align
289 // CHECK-NEXT:  [[ERROR:%.+]] = icmp ne i32 [[RET2]], 0
290 // CHECK-NEXT:  br i1 [[ERROR]], label %[[FAIL:.+]], label %[[END:[^,]+]]
291 //
292 // CHECK:       [[FAIL]]
293 // CHECK:       call void [[HVT6:@.+]]({{[^,]+}}, {{[^,]+}})
294 // CHECK:       br label {{%?}}[[END]]
295 // CHECK:       [[END]]
296 
297 
298 
299 
300 
301 
302 // Check that the offloading functions are emitted and that the parallel function
303 // is appropriately guarded.
304 
305 // CHECK:       define internal void [[HVT1]]([[S1]]* {{%.+}}, i[[SZ]] [[PARM1:%.+]], i[[SZ]] [[PARM2:%.+]])
306 // CHECK-DAG:   store i[[SZ]] [[PARM1]], i[[SZ]]* [[B_ADDR:%.+]], align
307 // CHECK-DAG:   store i[[SZ]] [[PARM2]], i[[SZ]]* [[CAPE_ADDR:%.+]], align
308 // CHECK-64:    [[CONVB:%.+]] = bitcast i[[SZ]]* [[B_ADDR]] to i32*
309 // CHECK:       [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPE_ADDR]] to i8*
310 // CHECK-64:    [[BV:%.+]] = load i32, i32* [[CONVB]], align
311 // CHECK-32:    [[BV:%.+]] = load i32, i32* [[B_ADDR]], align
312 // CHECK-64:    [[BC:%.+]] = bitcast i64* [[ARGA:%.+]] to i32*
313 // CHECK-64:    store i32 [[BV]], i32* [[BC]], align
314 // CHECK-64:    [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[ARGA]], align
315 // CHECK-32:    store i32 [[BV]], i32* [[ARGA:%.+]], align
316 // CHECK-32:    [[ARG:%.+]] = load i[[SZ]], i[[SZ]]* [[ARGA]], align
317 // CHECK:       [[IFC:%.+]] = load i8, i8* [[CONV]], align
318 // CHECK:       [[TB:%.+]] = trunc i8 [[IFC]] to i1
319 // CHECK:       br i1 [[TB]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]]
320 //
321 // CHECK:       [[IF_THEN]]
322 // CHECK:       call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%ident_t* [[DEF_LOC]], i32 2, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, [[S1]]*, i[[SZ]])* [[OMP_OUTLINED3:@.+]] to void (i32*, i32*, ...)*), [[S1]]* {{.+}}, i[[SZ]] [[ARG]])
323 // CHECK:       br label {{%?}}[[END:.+]]
324 //
325 // CHECK:       [[IF_ELSE]]
326 // CHECK:       call void @__kmpc_serialized_parallel(
327 // CHECK:       call void [[OMP_OUTLINED3]](i32* {{%.+}}, i32* {{%.+}}, [[S1]]* {{.+}}, i[[SZ]] [[ARG]])
328 // CHECK:       call void @__kmpc_end_serialized_parallel(
329 // CHECK:       br label {{%?}}[[END]]
330 //
331 // CHECK:       [[END]]
332 //
333 //
334 
335 
336 // CHECK:       define internal void [[HVT2]]([[S1]]* {{%.+}}, i[[SZ]] [[PARM:%.+]])
337 // CHECK-DAG:   store i[[SZ]] [[PARM]], i[[SZ]]* [[CAPE_ADDR:%.+]], align
338 // CHECK:       [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPE_ADDR]] to i8*
339 // CHECK:       [[IFC:%.+]] = load i8, i8* [[CONV]], align
340 // CHECK:       [[TB:%.+]] = trunc i8 [[IFC]] to i1
341 // CHECK:       br i1 [[TB]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]]
342 //
343 // CHECK:       [[IF_THEN]]
344 // CHECK:       call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%ident_t* [[DEF_LOC]], i32 1, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, [[S1]]*)* [[OMP_OUTLINED4:@.+]] to void (i32*, i32*, ...)*), [[S1]]* {{.+}})
345 // CHECK:       br label {{%?}}[[END:.+]]
346 //
347 // CHECK:       [[IF_ELSE]]
348 // CHECK:       call void @__kmpc_serialized_parallel(
349 // CHECK:       call void [[OMP_OUTLINED4]](i32* {{%.+}}, i32* {{%.+}}, [[S1]]* {{.+}})
350 // CHECK:       call void @__kmpc_end_serialized_parallel(
351 // CHECK:       br label {{%?}}[[END]]
352 //
353 // CHECK:       [[END]]
354 //
355 //
356 
357 
358 
359 
360 
361 
362 
363 
364 // CHECK:       define internal void [[HVT3]](i[[SZ]] [[PARM:%.+]])
365 // CHECK-DAG:   store i[[SZ]] [[PARM]], i[[SZ]]* [[CAPE_ADDR:%.+]], align
366 // CHECK:       [[CONV:%.+]] = bitcast i[[SZ]]* [[CAPE_ADDR]] to i8*
367 // CHECK:       [[IFC:%.+]] = load i8, i8* [[CONV]], align
368 // CHECK:       [[TB:%.+]] = trunc i8 [[IFC]] to i1
369 // CHECK:       br i1 [[TB]], label {{%?}}[[IF_THEN:.+]], label {{%?}}[[IF_ELSE:.+]]
370 //
371 // CHECK:       [[IF_THEN]]
372 // CHECK:       call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%ident_t* [[DEF_LOC]], i32 0, void (i32*, i32*, ...)* bitcast (void (i32*, i32*)* [[OMP_OUTLINED1:@.+]] to void (i32*, i32*, ...)*))
373 // CHECK:       br label {{%?}}[[END:.+]]
374 //
375 // CHECK:       [[IF_ELSE]]
376 // CHECK:       call void @__kmpc_serialized_parallel(
377 // CHECK:       call void [[OMP_OUTLINED1]](i32* {{%.+}}, i32* {{%.+}})
378 // CHECK:       call void @__kmpc_end_serialized_parallel(
379 // CHECK:       br label {{%?}}[[END]]
380 //
381 // CHECK:       [[END]]
382 //
383 //
384 // CHECK:       define internal void [[HVT4]]()
385 // CHECK:       call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%ident_t* [[DEF_LOC]], i32 0, void (i32*, i32*, ...)* bitcast (void (i32*, i32*)* [[OMP_OUTLINED2:@.+]] to void (i32*, i32*, ...)*))
386 // CHECK-NEXT:  ret
387 //
388 //
389 
390 
391 
392 
393 
394 // CHECK:       define internal void [[HVT5]](
395 // CHECK-NOT:   @__kmpc_fork_call
396 // CHECK:       call void @__kmpc_serialized_parallel(
397 // CHECK:       call void [[OMP_OUTLINED5:@.+]](i32* {{%.+}}, i32* {{%.+}}, i[[SZ]] {{.+}})
398 // CHECK:       call void @__kmpc_end_serialized_parallel(
399 // CHECK:       ret
400 //
401 //
402 
403 
404 // CHECK:       define internal void [[HVT6]](
405 // CHECK-NOT:   call void @__kmpc_serialized_parallel(
406 // CHECK-NOT:   call void [[OMP_OUTLINED5:@.+]](i32* {{%.+}}, i32* {{%.+}}, i[[SZ]] {{.+}})
407 // CHECK-NOT:   call void @__kmpc_end_serialized_parallel(
408 // CHECK:       call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call(%ident_t* [[DEF_LOC]], i32 2, void (i32*, i32*, ...)* bitcast (void (i32*, i32*, i[[SZ]], i[[SZ]])* [[OMP_OUTLINED5:@.+]] to void (i32*, i32*, ...)*),
409 // CHECK:       ret
410 //
411 //
412 
413 
414 
415 #endif
416