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 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