1 // RUN: %clang_cc1 -verify -fopenmp -x c -triple x86_64-unknown-linux -emit-llvm %s -o - | FileCheck %s --check-prefix HOST
2 // RUN: %clang_cc1 -fopenmp -x c -triple x86_64-unknown-linux -emit-pch -o %t -fopenmp-version=45 %s
3 // RUN: %clang_cc1 -fopenmp -x c -triple x86_64-unknown-linux -include-pch %t -verify %s -emit-llvm -o - -fopenmp-version=45 | FileCheck %s --check-prefix HOST
4 // 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 -fopenmp-version=45
5 // RUN: %clang_cc1 -verify -fopenmp -x c -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - -fopenmp-version=45 | FileCheck %s --check-prefix GPU
6 // RUN: %clang_cc1 -verify -fopenmp -x c -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -emit-pch -o %t -fopenmp-version=45
7 // RUN: %clang_cc1 -verify -fopenmp -x c -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -o - -fopenmp-version=45 | FileCheck %s --check-prefix GPU
8
9 // RUN: %clang_cc1 -verify -fopenmp -x c -triple x86_64-unknown-linux -emit-llvm %s -o - | FileCheck %s --check-prefix HOST
10 // RUN: %clang_cc1 -fopenmp -x c -triple x86_64-unknown-linux -emit-pch -o %t %s
11 // RUN: %clang_cc1 -fopenmp -x c -triple x86_64-unknown-linux -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix HOST
12 // 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
13 // RUN: %clang_cc1 -verify -fopenmp -x c -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s --check-prefix GPU
14 // RUN: %clang_cc1 -verify -fopenmp -x c -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -emit-pch -o %t
15 // RUN: %clang_cc1 -verify -fopenmp -x c -triple nvptx64-unknown-unknown -aux-triple powerpc64le-unknown-unknown -emit-llvm %s -fopenmp-is-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -o - | FileCheck %s --check-prefix GPU
16 // expected-no-diagnostics
17
18 #ifndef HEADER
19 #define HEADER
20
dev(double i)21 int dev(double i) { return 0; }
22
hst(double i)23 int hst(double i) { return 1; }
24
25 #pragma omp declare variant(hst) match(device = {kind(host)})
26 #pragma omp declare variant(dev) match(device = {kind(gpu)})
27 int base();
28
29 // HOST-LABEL: define{{.*}} void @foo()
30 // HOST: call i32 @hst(double -1.000000e+00)
31 // HOST: call i32 @hst(double -2.000000e+00)
32 // HOST: call void [[OFFL:@.+_foo_l36]]()
foo()33 void foo() {
34 base(-1);
35 hst(-2);
36 #pragma omp target
37 {
38 base(-3);
39 dev(-4);
40 }
41 }
42
43 // HOST: define {{.*}}void [[OFFL]]()
44 // HOST: call i32 @hst(double -3.000000e+00)
45 // HOST: call i32 @dev(double -4.000000e+00)
46
47 // GPU: define {{.*}}void @__omp_offloading_{{.+}}_foo_l36()
48 // GPU: call i32 @dev(double -3.000000e+00)
49 // GPU: call i32 @dev(double -4.000000e+00)
50
51 // GPU-NOT: @base
52 // GPU: define {{.*}}i32 @dev(double
53 // GPU: ret i32 0
54
55 #endif // HEADER
56