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|Base|virtual_}}
16 // CHECK-DAG: Bake
17 // CHECK-NOT: @{{hhh|ggg|fff|eee}} =
18 // CHECK-DAG: @aaa = external global i32,
19 // CHECK-DAG: @bbb ={{ dso_local | }}global i32 0,
20 // CHECK-DAG: weak constant %struct.__tgt_offload_entry { i8* bitcast (i32* @bbb to i8*),
21 // CHECK-DAG: @ccc = external global i32,
22 // CHECK-DAG: @ddd ={{ dso_local | }}global i32 0,
23 // CHECK-DAG: @hhh_decl_tgt_ref_ptr = weak global i32* null
24 // CHECK-DAG: @ggg_decl_tgt_ref_ptr = weak global i32* null
25 // CHECK-DAG: @fff_decl_tgt_ref_ptr = weak global i32* null
26 // CHECK-DAG: @eee_decl_tgt_ref_ptr = weak global i32* null
27 // CHECK-DAG: @{{.*}}maini1{{.*}}aaa = internal global i64 23,
28 // CHECK-DAG: @b ={{ dso_local | }}global i32 15,
29 // CHECK-DAG: @d ={{ dso_local | }}global i32 0,
30 // CHECK-DAG: @c = external global i32,
31 // CHECK-DAG: @globals ={{ dso_local | }}global %struct.S zeroinitializer,
32 // CHECK-DAG: [[STAT:@.+stat]] = internal global %struct.S zeroinitializer,
33 // CHECK-DAG: [[STAT_REF:@.+]] = internal constant %struct.S* [[STAT]]
34 // CHECK-DAG: @out_decl_target ={{ dso_local | }}global i32 0,
35 // CHECK-DAG: @llvm.used = appending global [2 x i8*] [i8* bitcast (void ()* @__omp_offloading__{{.+}}_globals_l[[@LINE+80]]_ctor to i8*), i8* bitcast (void ()* @__omp_offloading__{{.+}}_stat_l[[@LINE+81]]_ctor to i8*)],
36 // CHECK-DAG: @llvm.compiler.used = appending global [1 x i8*] [i8* bitcast (%struct.S** [[STAT_REF]] to i8*)],
37 
38 // CHECK-DAG: define {{.*}}i32 @{{.*}}{{foo|bar|baz2|baz3|FA|f_method}}{{.*}}()
39 // CHECK-DAG: define {{.*}}void @{{.*}}TemplateClass{{.*}}(%class.TemplateClass* %{{.*}})
40 // CHECK-DAG: define {{.*}}i32 @{{.*}}TemplateClass{{.*}}f_method{{.*}}(%class.TemplateClass* %{{.*}})
41 // CHECK-DAG: define {{.*}}void @__omp_offloading__{{.*}}_globals_l[[@LINE+74]]_ctor()
42 
43 #ifndef HEADER
44 #define HEADER
45 
46 #pragma omp declare target
47 extern int bbb;
48 #pragma omp end declare target
49 #pragma omp declare target
50 extern int bbb;
51 #pragma omp end declare target
52 
53 #pragma omp declare target
54 extern int aaa;
55 int bbb = 0;
56 extern int ccc;
57 int ddd = 0;
58 #pragma omp end declare target
59 
60 #pragma omp declare target
61 extern int bbb;
62 #pragma omp end declare target
63 
64 extern int eee;
65 int fff = 0;
66 extern int ggg;
67 int hhh = 0;
68 #pragma omp declare target link(eee, fff, ggg, hhh)
69 
70 int out_decl_target = 0;
71 #pragma omp declare target
72 void lambda () {
73 #ifdef __cpp_lambdas
74   (void)[&] { (void)out_decl_target; };
75 #else
76 #pragma clang __debug captured
77   {
78     (void)out_decl_target;
79   }
80 #endif
81 };
82 #pragma omp end declare target
83 
84 template <typename T>
85 class TemplateClass {
86   T a;
87 public:
88   TemplateClass() {}
89   T f_method() const { return a; }
90 };
91 
92 int foo();
93 
94 static int baz1() { return 0; }
95 
96 int baz2();
97 
98 int baz4() { return 5; }
99 
100 template <typename T>
101 T FA() {
102   TemplateClass<T> s;
103   return s.f_method();
104 }
105 
106 #pragma omp declare target
107 struct S {
108   int a;
109   S(int a) : a(a) {}
110 };
111 
112 int foo() { return 0; }
113 int b = 15;
114 int d;
115 S globals(d);
116 static S stat(d);
117 #pragma omp end declare target
118 int c;
119 
120 int bar() { return 1 + foo() + bar() + baz1() + baz2(); }
121 
122 int maini1() {
123   int a;
124   static long aa = 32 + bbb + ccc + fff + ggg;
125 // CHECK-DAG: define weak void @__omp_offloading_{{.*}}maini1{{.*}}_l[[@LINE+1]](i32* dereferenceable{{.*}}, i64 {{.*}}, i64 {{.*}})
126 #pragma omp target map(tofrom \
127                        : a, b)
128   {
129     S s(a);
130     static long aaa = 23;
131     a = foo() + bar() + b + c + d + aa + aaa + FA<int>();
132   }
133   return baz4();
134 }
135 
136 int baz3() { return 2 + baz2(); }
137 int baz2() {
138 // CHECK-DAG: define weak void @__omp_offloading_{{.*}}baz2{{.*}}_l[[@LINE+1]](i64 {{.*}})
139 #pragma omp target parallel
140   ++c;
141   return 2 + baz3();
142 }
143 
144 extern int create() throw();
145 
146 static __typeof(create) __t_create __attribute__((__weakref__("__create")));
147 
148 int baz5() {
149   bool a;
150 // CHECK-DAG: define weak void @__omp_offloading_{{.*}}baz5{{.*}}_l[[@LINE+1]](i64 {{.*}})
151 #pragma omp target
152   a = __extension__(void *) & __t_create != 0;
153   return a;
154 }
155 
156 template <typename T>
157 struct Base {
158   virtual ~Base() {}
159 };
160 
161 template class Base<int>;
162 
163 template <typename T>
164 struct Bake {
165   virtual ~Bake() {}
166 };
167 
168 #pragma omp declare target
169 template class Bake<int>;
170 #pragma omp end declare target
171 
172 struct BaseNonT {
173   virtual ~BaseNonT() {}
174 };
175 
176 #pragma omp declare target
177 struct BakeNonT {
178   virtual ~BakeNonT() {}
179 };
180 #pragma omp end declare target
181 
182 template <typename T>
183 struct B {
184   virtual void virtual_foo();
185 };
186 
187 void new_bar() { new B<int>(); }
188 
189 template <typename T>
190 void B<T>::virtual_foo() {
191 #pragma omp target
192   {}
193 }
194 
195 struct A {
196   virtual void emitted() {}
197 };
198 
199 template <typename T>
200 struct C : public A {
201   virtual void emitted();
202 };
203 
204 template <typename T>
205 void C<T>::emitted() {
206 #pragma omp target
207   {}
208 }
209 
210 int main() {
211   A *X = new C<int>();
212   X->emitted();
213   return 0;
214 }
215 
216 // CHECK-DAG: define {{.*}}void @__omp_offloading_{{.*}}virtual_foo{{.*}}_l[[@LINE-25]]()
217 // CHECK-DAG: define {{.*}}void @__omp_offloading_{{.*}}emitted{{.*}}_l[[@LINE-11]]()
218 
219 // CHECK-DAG: declare extern_weak signext i32 @__create()
220 
221 // CHECK-NOT: define {{.*}}{{baz1|baz4|maini1|Base|virtual_}}
222 
223 // CHECK-DAG: !{i32 1, !"aaa", i32 0, i32 {{[0-9]+}}}
224 // CHECK-DAG: !{i32 1, !"ccc", i32 0, i32 {{[0-9]+}}}
225 // CHECK-DAG: !{{{.+}}virtual_foo
226 
227 #endif // HEADER
228