{"owner":"nuudlman","repo":"llvm","channel":"main","channels":["main"],"can_delete":true,"inode":{"File":{"path":[{"basename":"clang","pos":"7OVOGA6QKF5OI.EJBFAJQ","meta":576},{"basename":"test","pos":"7OVOGA6QKF5OI.RXSO6JQ","meta":576},{"basename":"OpenMP","pos":"7OVOGA6QKF5OI.T5I3CJY","meta":576},{"basename":"nvptx_force_full_runtime_SPMD_codegen.cpp","pos":"7OVOGA6QKF5OI.5EI7AKI","meta":0}],"file":"// Test target codegen - host bc file has to be created first.\n// 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\n// 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 - -fopenmp-cuda-force-full-runtime | FileCheck %s\n// 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\n// 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 - -fopenmp-cuda-force-full-runtime | FileCheck %s\n// 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 - -fopenmp-cuda-force-full-runtime | FileCheck %s\n// expected-no-diagnostics\n#ifndef HEADER\n#define HEADER\n\n// CHECK-NOT: @__omp_offloading_{{.+}}_exec_mode = weak constant i8 1\n\nvoid foo() {\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target teams distribute parallel for simd\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for simd schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for simd schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for simd schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for simd schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for simd schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for simd schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\nint a;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target teams distribute parallel for lastprivate(a)\n for (int i = 0; i < 10; ++i)\n a = i;\n#pragma omp target teams distribute parallel for schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams distribute parallel for schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target teams\n#pragma omp distribute parallel for simd\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for simd schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for simd schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for simd schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for simd schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for simd schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for simd schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target teams\n#pragma omp distribute parallel for\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target teams\n#pragma omp distribute parallel for schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target\n#pragma omp teams\n#pragma omp distribute parallel for\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp teams\n#pragma omp distribute parallel for schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp teams\n#pragma omp distribute parallel for schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp teams\n#pragma omp distribute parallel for schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp teams\n#pragma omp distribute parallel for schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp teams\n#pragma omp distribute parallel for schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp teams\n#pragma omp distribute parallel for schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target parallel for\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel for schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel for schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel for schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel for schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel for schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel for schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target parallel\n#pragma omp for simd\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel\n#pragma omp for simd schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel\n#pragma omp for simd schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel\n#pragma omp for simd schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel\n#pragma omp for simd schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel\n#pragma omp for simd schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target parallel\n#pragma omp for simd schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target\n#pragma omp parallel\n#pragma omp for simd ordered\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel\n#pragma omp for simd schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel\n#pragma omp for simd schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel\n#pragma omp for simd schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel\n#pragma omp for simd schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel\n#pragma omp for simd schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel\n#pragma omp for simd schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n// CHECK: call i32 @__kmpc_target_init({{.*}}, i8 2, i1 false, i1 true)\n#pragma omp target\n#pragma omp parallel for\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel for schedule(static)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel for schedule(static, 1)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel for schedule(auto)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel for schedule(runtime)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel for schedule(dynamic)\n for (int i = 0; i < 10; ++i)\n ;\n#pragma omp target\n#pragma omp parallel for schedule(guided)\n for (int i = 0; i < 10; ++i)\n ;\n}\n\n#endif\n\n"}},"token":"Vp7KzfixS7eY1XpKG6+dSYiy5L49TFNxnpgiG8JDr7A=","login":null}