1 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
2 // RUN: %clang_cc1 -verify -fopenmp -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
3 // RUN: %clang_cc1 -verify -fopenmp -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 -emit-pch -o %t
4 // RUN: %clang_cc1 -verify -fopenmp -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 -include-pch %t -o - | FileCheck %s
5 
6 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc
7 // RUN: %clang_cc1 -verify -fopenmp-simd -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 SIMD-ONLY
8 // RUN: %clang_cc1 -verify -fopenmp-simd -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 -emit-pch -o %t
9 // RUN: %clang_cc1 -verify -fopenmp-simd -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 -include-pch %t -verify -o - | FileCheck %s --check-prefix SIMD-ONLY
10 
11 // expected-no-diagnostics
12 
13 // SIMD-ONLY-NOT: {{__kmpc|__tgt}}
14 
15 // CHECK-NOT: define {{.*}}{{baz1|baz4|maini1}}
16 // CHECK-DAG: @{{.*}}maini1{{.*}}aaa = internal global i64 23,
17 // CHECK-DAG: @b = global i32 15,
18 // CHECK-DAG: @d = global i32 0,
19 // CHECK-DAG: @c = external global i32,
20 // CHECK-DAG: @globals = global %struct.S zeroinitializer,
21 // CHECK-DAG: [[STAT:@.+stat]] = internal global %struct.S zeroinitializer,
22 // CHECK-DAG: [[STAT_REF:@.+]] = internal constant %struct.S* [[STAT]]
23 // CHECK-DAG: @out_decl_target = global i32 0,
24 // CHECK-DAG: @llvm.used = appending global [2 x i8*] [i8* bitcast (void ()* @__omp_offloading__{{.+}}_globals_l[[@LINE+56]]_ctor to i8*), i8* bitcast (void ()* @__omp_offloading__{{.+}}_stat_l[[@LINE+57]]_ctor to i8*)],
25 // CHECK-DAG: @llvm.compiler.used = appending global [1 x i8*] [i8* bitcast (%struct.S** [[STAT_REF]] to i8*)],
26 
27 // CHECK-DAG: define {{.*}}i32 @{{.*}}{{foo|bar|baz2|baz3|FA|f_method}}{{.*}}()
28 // CHECK-DAG: define {{.*}}void @{{.*}}TemplateClass{{.*}}(%class.TemplateClass* %{{.*}})
29 // CHECK-DAG: define {{.*}}i32 @{{.*}}TemplateClass{{.*}}f_method{{.*}}(%class.TemplateClass* %{{.*}})
30 // CHECK-DAG: define {{.*}}void @__omp_offloading__{{.*}}_globals_l[[@LINE+50]]_ctor()
31 
32 #ifndef HEADER
33 #define HEADER
34 
35 int out_decl_target = 0;
36 #pragma omp declare target
37 void lambda () {
38 #ifdef __cpp_lambdas
39   (void)[&] { (void)out_decl_target; };
40 #else
41 #pragma clang __debug captured
42   {
43     (void)out_decl_target;
44   }
45 #endif
46 };
47 #pragma omp end declare target
48 
49 template <typename T>
50 class TemplateClass {
51   T a;
52 public:
53   TemplateClass() {}
54   T f_method() const { return a; }
55 };
56 
57 int foo();
58 
59 int baz1();
60 
61 int baz2();
62 
63 int baz4() { return 5; }
64 
65 template <typename T>
66 T FA() {
67   TemplateClass<T> s;
68   return s.f_method();
69 }
70 
71 #pragma omp declare target
72 struct S {
73   int a;
74   S(int a) : a(a) {}
75 };
76 
77 int foo() { return 0; }
78 int b = 15;
79 int d;
80 S globals(d);
81 static S stat(d);
82 #pragma omp end declare target
83 int c;
84 
85 int bar() { return 1 + foo() + bar() + baz1() + baz2(); }
86 
87 int maini1() {
88   int a;
89   static long aa = 32;
90 // CHECK-DAG: define weak void @__omp_offloading_{{.*}}maini1{{.*}}_l[[@LINE+1]](i32* dereferenceable{{.*}}, i64 {{.*}}, i64 {{.*}})
91 #pragma omp target map(tofrom \
92                        : a, b)
93   {
94     S s(a);
95     static long aaa = 23;
96     a = foo() + bar() + b + c + d + aa + aaa + FA<int>();
97   }
98   return baz4();
99 }
100 
101 int baz3() { return 2 + baz2(); }
102 int baz2() {
103 // CHECK-DAG: define weak void @__omp_offloading_{{.*}}baz2{{.*}}_l[[@LINE+1]](i64 {{.*}})
104 #pragma omp target parallel
105   ++c;
106   return 2 + baz3();
107 }
108 
109 // CHECK-NOT: define {{.*}}{{baz1|baz4|maini1}}
110 #endif // HEADER
111