1 // Test target codegen - host bc file has to be created first.
2 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm-bc %s -o %t-ppc-host.bc
3 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx64-unknown-unknown -fopenmp-targets=nvptx64-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s
4 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=nvptx-nvidia-cuda -emit-llvm-bc %s -o %t-x86-host.bc
5 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple nvptx-unknown-unknown -fopenmp-targets=nvptx-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck %s
6 // RUN: %clang_cc1 -verify -fopenmp -fexceptions -fcxx-exceptions -x c++ -triple nvptx-unknown-unknown -fopenmp-targets=nvptx-nvidia-cuda -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck %s
7 // expected-no-diagnostics
8 #ifndef HEADER
9 #define HEADER
10
11 int a;
12
13 // CHECK-NOT: @__omp_offloading_{{.+}}_exec_mode = weak constant i8 1
14 // CHECK-DAG: [[DISTR_LIGHT:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 2050, i32 3, i32 0, i8* getelementptr inbounds
15 // CHECK-DAG: [[FOR_LIGHT:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 514, i32 3, i32 0, i8* getelementptr inbounds
16 // CHECK-DAG: [[LIGHT:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 2, i32 3, i32 0, i8* getelementptr inbounds
17 // CHECK-DAG: [[DISTR_FULL:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 2050, i32 1, i32 0, i8* getelementptr inbounds
18 // CHECK-DAG: [[FULL:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 2, i32 1, i32 0, i8* getelementptr inbounds
19 // CHECK-DAG: [[BAR_LIGHT:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 66, i32 3, i32 0, i8* getelementptr inbounds
20 // CHECK-DAG: [[BAR_FULL:@.+]] = private unnamed_addr constant %struct.ident_t { i32 0, i32 66, i32 1, i32 0, i8* getelementptr inbounds
21 // CHECK-NOT: @__omp_offloading_{{.+}}_exec_mode = weak constant i8 1
22
foo()23 void foo() {
24 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
25 // CHECK-DAG: [[DISTR_LIGHT]]
26 // CHECK-DAG: [[FOR_LIGHT]]
27 // CHECK-DAG: [[LIGHT]]
28 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
29 // CHECK-DAG: [[DISTR_LIGHT]]
30 // CHECK-DAG: [[FOR_LIGHT]]
31 // CHECK-DAG: [[LIGHT]]
32 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
33 // CHECK-DAG: [[DISTR_LIGHT]]
34 // CHECK-DAG: [[FOR_LIGHT]]
35 // CHECK-DAG: [[LIGHT]]
36 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
37 // CHECK-DAG: [[DISTR_FULL]]
38 // CHECK-DAG: [[FULL]]
39 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
40 // CHECK-DAG: [[DISTR_FULL]]
41 // CHECK-DAG: [[FULL]]
42 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
43 // CHECK-DAG: [[DISTR_FULL]]
44 // CHECK-DAG: [[FULL]]
45 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
46 // CHECK-DAG: [[DISTR_FULL]]
47 // CHECK-DAG: [[FULL]]
48 #pragma omp target teams distribute parallel for simd if(a)
49 for (int i = 0; i < 10; ++i)
50 ;
51 #pragma omp target teams distribute parallel for simd schedule(static)
52 for (int i = 0; i < 10; ++i)
53 ;
54 #pragma omp target teams distribute parallel for simd schedule(static, 1)
55 for (int i = 0; i < 10; ++i)
56 ;
57 #pragma omp target teams distribute parallel for simd schedule(auto)
58 for (int i = 0; i < 10; ++i)
59 ;
60 #pragma omp target teams distribute parallel for simd schedule(runtime)
61 for (int i = 0; i < 10; ++i)
62 ;
63 #pragma omp target teams distribute parallel for simd schedule(dynamic)
64 for (int i = 0; i < 10; ++i)
65 ;
66 #pragma omp target teams distribute parallel for simd schedule(guided)
67 for (int i = 0; i < 10; ++i)
68 ;
69 int a;
70 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
71 // CHECK-DAG: [[DISTR_LIGHT]]
72 // CHECK-DAG: [[FOR_LIGHT]]
73 // CHECK-DAG: [[LIGHT]]
74 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
75 // CHECK-DAG: [[DISTR_LIGHT]]
76 // CHECK-DAG: [[FOR_LIGHT]]
77 // CHECK-DAG: [[LIGHT]]
78 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
79 // CHECK-DAG: [[DISTR_LIGHT]]
80 // CHECK-DAG: [[FOR_LIGHT]]
81 // CHECK-DAG: [[LIGHT]]
82 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
83 // CHECK-DAG: [[DISTR_FULL]]
84 // CHECK-DAG: [[FULL]]
85 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
86 // CHECK-DAG: [[DISTR_FULL]]
87 // CHECK-DAG: [[FULL]]
88 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
89 // CHECK-DAG: [[DISTR_FULL]]
90 // CHECK-DAG: [[FULL]]
91 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
92 // CHECK-DAG: [[DISTR_FULL]]
93 // CHECK-DAG: [[FULL]]
94 #pragma omp target teams distribute parallel for lastprivate(a)
95 for (int i = 0; i < 10; ++i)
96 a = i;
97 #pragma omp target teams distribute parallel for schedule(static)
98 for (int i = 0; i < 10; ++i)
99 ;
100 #pragma omp target teams distribute parallel for schedule(static, 1)
101 for (int i = 0; i < 10; ++i)
102 ;
103 #pragma omp target teams distribute parallel for schedule(auto)
104 for (int i = 0; i < 10; ++i)
105 ;
106 #pragma omp target teams distribute parallel for schedule(runtime)
107 for (int i = 0; i < 10; ++i)
108 ;
109 #pragma omp target teams distribute parallel for schedule(dynamic)
110 for (int i = 0; i < 10; ++i)
111 ;
112 #pragma omp target teams distribute parallel for schedule(guided)
113 for (int i = 0; i < 10; ++i)
114 ;
115 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
116 // CHECK-DAG: [[DISTR_LIGHT]]
117 // CHECK-DAG: [[FOR_LIGHT]]
118 // CHECK-DAG: [[LIGHT]]
119 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
120 // CHECK-DAG: [[DISTR_LIGHT]]
121 // CHECK-DAG: [[FOR_LIGHT]]
122 // CHECK-DAG: [[LIGHT]]
123 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
124 // CHECK-DAG: [[DISTR_LIGHT]]
125 // CHECK-DAG: [[FOR_LIGHT]]
126 // CHECK-DAG: [[LIGHT]]
127 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
128 // CHECK-DAG: [[DISTR_FULL]]
129 // CHECK-DAG: [[FULL]]
130 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
131 // CHECK-DAG: [[DISTR_FULL]]
132 // CHECK-DAG: [[FULL]]
133 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
134 // CHECK-DAG: [[DISTR_FULL]]
135 // CHECK-DAG: [[FULL]]
136 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
137 // CHECK-DAG: [[DISTR_FULL]]
138 // CHECK-DAG: [[FULL]]
139 #pragma omp target teams
140 {
141 int b;
142 #pragma omp distribute parallel for simd
143 for (int i = 0; i < 10; ++i)
144 ;
145 ;
146 }
147 #pragma omp target teams
148 {
149 int b[] = {2, 3, sizeof(int)};
150 #pragma omp distribute parallel for simd schedule(static)
151 for (int i = 0; i < 10; ++i)
152 ;
153 }
154 #pragma omp target teams
155 {
156 int b;
157 #pragma omp distribute parallel for simd schedule(static, 1)
158 for (int i = 0; i < 10; ++i)
159 ;
160 int &c = b;
161 }
162 #pragma omp target teams
163 #pragma omp distribute parallel for simd schedule(auto)
164 for (int i = 0; i < 10; ++i)
165 ;
166 #pragma omp target teams
167 #pragma omp distribute parallel for simd schedule(runtime)
168 for (int i = 0; i < 10; ++i)
169 ;
170 #pragma omp target teams
171 #pragma omp distribute parallel for simd schedule(dynamic)
172 for (int i = 0; i < 10; ++i)
173 ;
174 #pragma omp target teams
175 #pragma omp distribute parallel for simd schedule(guided)
176 for (int i = 0; i < 10; ++i)
177 ;
178 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
179 // CHECK-DAG: [[DISTR_LIGHT]]
180 // CHECK-DAG: [[FOR_LIGHT]]
181 // CHECK-DAG: [[LIGHT]]
182 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
183 // CHECK-DAG: [[DISTR_LIGHT]]
184 // CHECK-DAG: [[FOR_LIGHT]]
185 // CHECK-DAG: [[LIGHT]]
186 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
187 // CHECK-DAG: [[DISTR_LIGHT]]
188 // CHECK-DAG: [[FOR_LIGHT]]
189 // CHECK-DAG: [[LIGHT]]
190 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
191 // CHECK-DAG: [[DISTR_FULL]]
192 // CHECK-DAG: [[FULL]]
193 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
194 // CHECK-DAG: [[DISTR_FULL]]
195 // CHECK-DAG: [[FULL]]
196 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
197 // CHECK-DAG: [[DISTR_FULL]]
198 // CHECK-DAG: [[FULL]]
199 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
200 // CHECK-DAG: [[DISTR_FULL]]
201 // CHECK-DAG: [[FULL]]
202 #pragma omp target teams
203 #pragma omp distribute parallel for
204 for (int i = 0; i < 10; ++i)
205 ;
206 #pragma omp target teams
207 #pragma omp distribute parallel for schedule(static)
208 for (int i = 0; i < 10; ++i)
209 ;
210 #pragma omp target teams
211 #pragma omp distribute parallel for schedule(static, 1)
212 for (int i = 0; i < 10; ++i)
213 ;
214 #pragma omp target teams
215 #pragma omp distribute parallel for schedule(auto)
216 for (int i = 0; i < 10; ++i)
217 ;
218 #pragma omp target teams
219 #pragma omp distribute parallel for schedule(runtime)
220 for (int i = 0; i < 10; ++i)
221 ;
222 #pragma omp target teams
223 #pragma omp distribute parallel for schedule(dynamic)
224 for (int i = 0; i < 10; ++i)
225 ;
226 #pragma omp target teams
227 #pragma omp distribute parallel for schedule(guided)
228 for (int i = 0; i < 10; ++i)
229 ;
230 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
231 // CHECK-DAG: [[DISTR_LIGHT]]
232 // CHECK-DAG: [[FOR_LIGHT]]
233 // CHECK-DAG: [[LIGHT]]
234 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
235 // CHECK-DAG: [[DISTR_LIGHT]]
236 // CHECK-DAG: [[FOR_LIGHT]]
237 // CHECK-DAG: [[LIGHT]]
238 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
239 // CHECK-DAG: [[DISTR_LIGHT]]
240 // CHECK-DAG: [[FOR_LIGHT]]
241 // CHECK-DAG: [[LIGHT]]
242 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
243 // CHECK-DAG: [[DISTR_FULL]]
244 // CHECK-DAG: [[FULL]]
245 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
246 // CHECK-DAG: [[DISTR_FULL]]
247 // CHECK-DAG: [[FULL]]
248 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
249 // CHECK-DAG: [[DISTR_FULL]]
250 // CHECK-DAG: [[FULL]]
251 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
252 // CHECK-DAG: [[DISTR_FULL]]
253 // CHECK-DAG: [[FULL]]
254 #pragma omp target
255 #pragma omp teams
256 #pragma omp distribute parallel for
257 for (int i = 0; i < 10; ++i)
258 ;
259 #pragma omp target
260 #pragma omp teams
261 #pragma omp distribute parallel for schedule(static)
262 for (int i = 0; i < 10; ++i)
263 ;
264 #pragma omp target
265 #pragma omp teams
266 #pragma omp distribute parallel for schedule(static, 1)
267 for (int i = 0; i < 10; ++i)
268 ;
269 #pragma omp target
270 #pragma omp teams
271 #pragma omp distribute parallel for schedule(auto)
272 for (int i = 0; i < 10; ++i)
273 ;
274 #pragma omp target
275 #pragma omp teams
276 #pragma omp distribute parallel for schedule(runtime)
277 for (int i = 0; i < 10; ++i)
278 ;
279 #pragma omp target
280 #pragma omp teams
281 #pragma omp distribute parallel for schedule(dynamic)
282 for (int i = 0; i < 10; ++i)
283 ;
284 #pragma omp target
285 #pragma omp teams
286 #pragma omp distribute parallel for schedule(guided)
287 for (int i = 0; i < 10; ++i)
288 ;
289 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
290 // CHECK-DAG: [[FOR_LIGHT]]
291 // CHECK-DAG: [[LIGHT]]
292 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
293 // CHECK-DAG: [[FOR_LIGHT]]
294 // CHECK-DAG: [[LIGHT]]
295 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
296 // CHECK-DAG: [[FOR_LIGHT]]
297 // CHECK-DAG: [[LIGHT]]
298 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
299 // CHECK-DAG: [[FULL]]
300 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
301 // CHECK-DAG: [[FULL]]
302 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
303 // CHECK-DAG: [[FULL]]
304 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
305 // CHECK-DAG: [[FULL]]
306 #pragma omp target parallel for if(a)
307 for (int i = 0; i < 10; ++i)
308 ;
309 #pragma omp target parallel for schedule(static)
310 for (int i = 0; i < 10; ++i)
311 ;
312 #pragma omp target parallel for schedule(static, 1)
313 for (int i = 0; i < 10; ++i)
314 ;
315 #pragma omp target parallel for schedule(auto)
316 for (int i = 0; i < 10; ++i)
317 ;
318 #pragma omp target parallel for schedule(runtime)
319 for (int i = 0; i < 10; ++i)
320 ;
321 #pragma omp target parallel for schedule(dynamic)
322 for (int i = 0; i < 10; ++i)
323 ;
324 #pragma omp target parallel for schedule(guided)
325 for (int i = 0; i < 10; ++i)
326 ;
327 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
328 // CHECK-DAG: [[FOR_LIGHT]]
329 // CHECK-DAG: [[LIGHT]]
330 // CHECK-DAG: [[BAR_LIGHT]]
331 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
332 // CHECK-DAG: [[FOR_LIGHT]]
333 // CHECK-DAG: [[LIGHT]]
334 // CHECK-DAG: [[BAR_LIGHT]]
335 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
336 // CHECK-DAG: [[FOR_LIGHT]]
337 // CHECK-DAG: [[LIGHT]]
338 // CHECK-DAG: [[BAR_LIGHT]]
339 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
340 // CHECK-DAG: [[FULL]]
341 // CHECK-DAG: [[BAR_FULL]]
342 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
343 // CHECK-DAG: [[FULL]]
344 // CHECK-DAG: [[BAR_FULL]]
345 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
346 // CHECK-DAG: [[FULL]]
347 // CHECK-DAG: [[BAR_FULL]]
348 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
349 // CHECK-DAG: [[FULL]]
350 // CHECK-DAG: [[BAR_FULL]]
351 #pragma omp target parallel if(a)
352 #pragma omp for simd
353 for (int i = 0; i < 10; ++i)
354 ;
355 #pragma omp target parallel
356 #pragma omp for simd schedule(static)
357 for (int i = 0; i < 10; ++i)
358 ;
359 #pragma omp target parallel
360 #pragma omp for simd schedule(static, 1)
361 for (int i = 0; i < 10; ++i)
362 ;
363 #pragma omp target parallel
364 #pragma omp for simd schedule(auto)
365 for (int i = 0; i < 10; ++i)
366 ;
367 #pragma omp target parallel
368 #pragma omp for simd schedule(runtime)
369 for (int i = 0; i < 10; ++i)
370 ;
371 #pragma omp target parallel
372 #pragma omp for simd schedule(dynamic)
373 for (int i = 0; i < 10; ++i)
374 ;
375 #pragma omp target parallel
376 #pragma omp for simd schedule(guided)
377 for (int i = 0; i < 10; ++i)
378 ;
379 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
380 // CHECK-DAG: [[FULL]]
381 // CHECK-DAG: [[BAR_FULL]]
382 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
383 // CHECK-DAG: [[FOR_LIGHT]]
384 // CHECK-DAG: [[LIGHT]]
385 // CHECK-DAG: [[BAR_LIGHT]]
386 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
387 // CHECK-DAG: [[FOR_LIGHT]]
388 // CHECK-DAG: [[LIGHT]]
389 // CHECK-DAG: [[BAR_LIGHT]]
390 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
391 // CHECK-DAG: [[FULL]]
392 // CHECK-DAG: [[BAR_FULL]]
393 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
394 // CHECK-DAG: [[FULL]]
395 // CHECK-DAG: [[BAR_FULL]]
396 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
397 // CHECK-DAG: [[FULL]]
398 // CHECK-DAG: [[BAR_FULL]]
399 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
400 // CHECK-DAG: [[FULL]]
401 // CHECK-DAG: [[BAR_FULL]]
402 #pragma omp target
403 #pragma omp parallel
404 #pragma omp for simd ordered
405 for (int i = 0; i < 10; ++i)
406 ;
407 #pragma omp target
408 #pragma omp parallel
409 #pragma omp for simd schedule(static)
410 for (int i = 0; i < 10; ++i)
411 ;
412 #pragma omp target
413 #pragma omp parallel
414 #pragma omp for simd schedule(static, 1)
415 for (int i = 0; i < 10; ++i)
416 ;
417 #pragma omp target
418 #pragma omp parallel
419 #pragma omp for simd schedule(auto)
420 for (int i = 0; i < 10; ++i)
421 ;
422 #pragma omp target
423 #pragma omp parallel
424 #pragma omp for simd schedule(runtime)
425 for (int i = 0; i < 10; ++i)
426 ;
427 #pragma omp target
428 #pragma omp parallel
429 #pragma omp for simd schedule(dynamic)
430 for (int i = 0; i < 10; ++i)
431 ;
432 #pragma omp target
433 #pragma omp parallel
434 #pragma omp for simd schedule(guided)
435 for (int i = 0; i < 10; ++i)
436 ;
437 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
438 // CHECK-DAG: [[FOR_LIGHT]]
439 // CHECK-DAG: [[LIGHT]]
440 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
441 // CHECK-DAG: [[FOR_LIGHT]]
442 // CHECK-DAG: [[LIGHT]]
443 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 0)
444 // CHECK-DAG: [[FOR_LIGHT]]
445 // CHECK-DAG: [[LIGHT]]
446 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
447 // CHECK-DAG: [[FULL]]
448 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
449 // CHECK-DAG: [[FULL]]
450 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
451 // CHECK-DAG: [[FULL]]
452 // CHECK: call void @__kmpc_spmd_kernel_init(i32 {{.+}}, i16 1)
453 // CHECK-DAG: [[FULL]]
454 #pragma omp target
455 #pragma omp parallel for
456 for (int i = 0; i < 10; ++i)
457 ;
458 #pragma omp target
459 #pragma omp parallel for schedule(static)
460 for (int i = 0; i < 10; ++i)
461 ;
462 #pragma omp target
463 #pragma omp parallel for schedule(static, 1)
464 for (int i = 0; i < 10; ++i)
465 ;
466 #pragma omp target
467 #pragma omp parallel for schedule(auto)
468 for (int i = 0; i < 10; ++i)
469 ;
470 #pragma omp target
471 #pragma omp parallel for schedule(runtime)
472 for (int i = 0; i < 10; ++i)
473 ;
474 #pragma omp target
475 #pragma omp parallel for schedule(dynamic)
476 for (int i = 0; i < 10; ++i)
477 ;
478 #pragma omp target
479 #pragma omp parallel for schedule(guided)
480 for (int i = 0; i < 10; ++i)
481 ;
482 }
483
484 #endif
485
486