1 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -ast-print %s -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP45
2 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping
3 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP45
4 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=50 -DOMP5 -ast-print %s -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP50
5 // RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -DOMP5 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping
6 // RUN: %clang_cc1 -fopenmp -fopenmp-version=50 -DOMP5 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP50
7 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=51 -ast-print %s -Wno-openmp-mapping -DOMP51 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP51
8 // RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping -DOMP51
9 // RUN: %clang_cc1 -fopenmp -fopenmp-version=51 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping -DOMP51 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP51
10 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=52 -ast-print %s -Wno-openmp-mapping -DOMP52 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP52
11 // RUN: %clang_cc1 -fopenmp -fopenmp-version=52 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping -DOMP52
12 // RUN: %clang_cc1 -fopenmp -fopenmp-version=52 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping -DOMP52 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP52
13
14 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -ast-print %s -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP45
15 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping
16 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP45
17 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 -DOMP5 -ast-print %s -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP50
18 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -DOMP5 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping
19 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -DOMP5 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP50
20 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=51 -ast-print %s -Wno-openmp-mapping -DOMP51 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP51
21 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping -DOMP51
22 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=51 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping -DOMP51 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP51
23 // expected-no-diagnostics
24 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=52 -ast-print %s -Wno-openmp-mapping -DOMP52 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP52
25 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=52 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping -DOMP52
26 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=52 -std=c++11 -include-pch %t -verify %s -ast-print -Wno-openmp-mapping -DOMP52 | FileCheck %s --check-prefix=CHECK --check-prefix=OMP52
27 // expected-no-diagnostics
28
29 #ifndef HEADER
30 #define HEADER
31
32 typedef void **omp_allocator_handle_t;
33 extern const omp_allocator_handle_t omp_null_allocator;
34 extern const omp_allocator_handle_t omp_default_mem_alloc;
35 extern const omp_allocator_handle_t omp_large_cap_mem_alloc;
36 extern const omp_allocator_handle_t omp_const_mem_alloc;
37 extern const omp_allocator_handle_t omp_high_bw_mem_alloc;
38 extern const omp_allocator_handle_t omp_low_lat_mem_alloc;
39 extern const omp_allocator_handle_t omp_cgroup_mem_alloc;
40 extern const omp_allocator_handle_t omp_pteam_mem_alloc;
41 extern const omp_allocator_handle_t omp_thread_mem_alloc;
42
foo()43 void foo() {}
44
45 struct S {
SS46 S(): a(0) {}
SS47 S(int v) : a(v) {}
48 int a;
49 typedef int type;
50 };
51
52 template <typename T>
53 class S7 : public T {
54 protected:
55 T a;
S7()56 S7() : a(0) {}
57
58 public:
S7(typename T::type v)59 S7(typename T::type v) : a(v) {
60 #pragma omp target simd private(a) private(this->a) private(T::a)
61 for (int k = 0; k < a.a; ++k)
62 ++this->a.a;
63 }
operator =(S7 & s)64 S7 &operator=(S7 &s) {
65 #pragma omp target simd private(a) private(this->a)
66 for (int k = 0; k < s.a.a; ++k)
67 ++s.a.a;
68 return *this;
69 }
70 };
71
72 // CHECK: #pragma omp target simd private(this->a) private(this->a) private(T::a){{$}}
73 // CHECK: #pragma omp target simd private(this->a) private(this->a)
74 // CHECK: #pragma omp target simd private(this->a) private(this->a) private(this->S::a)
75
76 class S8 : public S7<S> {
S8()77 S8() {}
78
79 public:
S8(int v)80 S8(int v) : S7<S>(v){
81 #pragma omp target simd private(a) private(this->a) private(S7<S>::a)
82 for (int k = 0; k < a.a; ++k)
83 ++this->a.a;
84 }
operator =(S8 & s)85 S8 &operator=(S8 &s) {
86 #pragma omp target simd private(a) private(this->a)
87 for (int k = 0; k < s.a.a; ++k)
88 ++s.a.a;
89 return *this;
90 }
91 };
92
93 // CHECK: #pragma omp target simd private(this->a) private(this->a) private(this->S7<S>::a)
94 // CHECK: #pragma omp target simd private(this->a) private(this->a)
95
96 template <class T, int N>
tmain(T argc,T * argv)97 T tmain(T argc, T *argv) {
98 T b = argc, c, d, e, f, h;
99 T arr[N][10], arr1[N];
100 T i, j;
101 T s;
102 static T a;
103 // CHECK: static T a;
104 static T g;
105 const T clen = 5;
106 // CHECK: T clen = 5;
107 T *p;
108 #pragma omp threadprivate(g)
109 #pragma omp target simd linear(a) allocate(a)
110 // CHECK: #pragma omp target simd linear(a) allocate(a)
111 for (T i = 0; i < 2; ++i)
112 a = 2;
113 // CHECK-NEXT: for (T i = 0; i < 2; ++i)
114 // CHECK-NEXT: a = 2;
115 #pragma omp target simd allocate(f) private(argc, b), firstprivate(c, d), lastprivate(d, f) collapse(N) if (target :argc) reduction(+ : h)
116 for (int i = 0; i < 2; ++i)
117 for (int j = 0; j < 2; ++j)
118 for (int j = 0; j < 2; ++j)
119 for (int j = 0; j < 2; ++j)
120 for (int j = 0; j < 2; ++j)
121 foo();
122 // CHECK-NEXT: #pragma omp target simd allocate(f) private(argc,b) firstprivate(c,d) lastprivate(d,f) collapse(N) if(target: argc) reduction(+: h)
123 // CHECK-NEXT: for (int i = 0; i < 2; ++i)
124 // CHECK-NEXT: for (int j = 0; j < 2; ++j)
125 // CHECK-NEXT: for (int j = 0; j < 2; ++j)
126 // CHECK-NEXT: for (int j = 0; j < 2; ++j)
127 // CHECK-NEXT: for (int j = 0; j < 2; ++j)
128 // CHECK-NEXT: foo();
129
130 #pragma omp target simd private(argc,b) firstprivate(argv) if(target:argc > 0) reduction(+:c, arr1[argc]) reduction(max:e, arr[:N][0:10])
131 for (T i = 0; i < 2; ++i) {}
132 // CHECK: #pragma omp target simd private(argc,b) firstprivate(argv) if(target: argc > 0) reduction(+: c,arr1[argc]) reduction(max: e,arr[:N][0:10])
133 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
134 // CHECK-NEXT: }
135
136 #pragma omp target simd if(N) reduction(^:e, f, arr[0:N][:argc]) reduction(&& : h)
137 for (T i = 0; i < 2; ++i) {}
138 // CHECK: #pragma omp target simd if(N) reduction(^: e,f,arr[0:N][:argc]) reduction(&&: h)
139 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
140 // CHECK-NEXT: }
141
142 #ifdef OMP5
143 #pragma omp target simd if(target:argc > 0) if (simd:argc) allocate(omp_default_mem_alloc:f) private(f) uses_allocators(omp_default_mem_alloc)
144 #else
145 #pragma omp target simd if(target:argc > 0)
146 #endif // OMP5
147 for (T i = 0; i < 2; ++i) {}
148 // OMP45: #pragma omp target simd if(target: argc > 0)
149 // OMP50: #pragma omp target simd if(target: argc > 0) if(simd: argc) allocate(omp_default_mem_alloc: f) private(f) uses_allocators(omp_default_mem_alloc)
150 // OMP51: #pragma omp target simd if(target: argc > 0)
151 // OMP52: #pragma omp target simd if(target: argc > 0)
152 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
153 // CHECK-NEXT: }
154
155 #pragma omp target simd if(N)
156 for (T i = 0; i < 2; ++i) {}
157 // CHECK: #pragma omp target simd if(N)
158 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
159 // CHECK-NEXT: }
160
161 #pragma omp target simd map(i)
162 for (T i = 0; i < 2; ++i) {}
163 // CHECK: #pragma omp target simd map(tofrom: i)
164 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
165 // CHECK-NEXT: }
166
167 #pragma omp target simd map(arr1[0:10], i)
168 for (T i = 0; i < 2; ++i) {}
169 // CHECK: #pragma omp target simd map(tofrom: arr1[0:10],i)
170 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
171 // CHECK-NEXT: }
172
173 #pragma omp target simd map(to: i) map(from: j)
174 for (T i = 0; i < 2; ++i) {}
175 // CHECK: #pragma omp target simd map(to: i) map(from: j)
176 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
177 // CHECK-NEXT: }
178
179 #pragma omp target simd map(always,alloc: i)
180 for (T i = 0; i < 2; ++i) {}
181 // CHECK: #pragma omp target simd map(always,alloc: i)
182 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
183 // CHECK-NEXT: }
184
185 #pragma omp target simd nowait
186 for (T i = 0; i < 2; ++i) {}
187 // CHECK: #pragma omp target simd nowait
188 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
189 // CHECK-NEXT: }
190
191 #pragma omp target simd depend(in : argc, arr[i:argc], arr1[:])
192 for (T i = 0; i < 2; ++i) {}
193 // CHECK: #pragma omp target simd depend(in : argc,arr[i:argc],arr1[:])
194 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
195 // CHECK-NEXT: }
196
197 #pragma omp target simd defaultmap(tofrom: scalar)
198 for (T i = 0; i < 2; ++i) {}
199 // CHECK: #pragma omp target simd defaultmap(tofrom: scalar)
200 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
201 // CHECK-NEXT: }
202
203 #pragma omp target simd safelen(clen-1)
204 for (T i = 0; i < 2; ++i) {}
205 // CHECK: #pragma omp target simd safelen(clen - 1)
206 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
207 // CHECK-NEXT: }
208
209 #pragma omp target simd simdlen(clen-1)
210 for (T i = 0; i < 2; ++i) {}
211 // CHECK: #pragma omp target simd simdlen(clen - 1)
212 // CHECK-NEXT: for (T i = 0; i < 2; ++i) {
213 // CHECK-NEXT: }
214
215 #pragma omp target simd aligned(arr1:N-1)
216 for (T i = 0; i < N; ++i) {}
217 // CHECK: #pragma omp target simd aligned(arr1: N - 1)
218 // CHECK-NEXT: for (T i = 0; i < N; ++i) {
219 // CHECK-NEXT: }
220
221 #pragma omp target simd is_device_ptr(p)
222 for (T i = 0; i < N; ++i) {}
223 // CHECK: #pragma omp target simd is_device_ptr(p)
224 // CHECK-NEXT: for (T i = 0; i < N; ++i) {
225 // CHECK-NEXT: }
226
227 return T();
228 }
229
main(int argc,char ** argv)230 int main(int argc, char **argv) {
231 int b = argc, c, d, e, f, h;
232 int arr[5][10], arr1[5];
233 int i, j;
234 int s;
235 static int a;
236 // CHECK: static int a;
237 const int clen = 5;
238 // CHECK: int clen = 5;
239 static float g;
240 #pragma omp threadprivate(g)
241 int *p;
242 #pragma omp target simd linear(a)
243 // CHECK: #pragma omp target simd linear(a)
244 for (int i = 0; i < 2; ++i)
245 a = 2;
246 // CHECK-NEXT: for (int i = 0; i < 2; ++i)
247 // CHECK-NEXT: a = 2;
248
249 #ifdef OMP52
250 #pragma omp target simd private(argc, b), firstprivate(argv, c), lastprivate(d, f) collapse(2) if (target: argc) reduction(+ : h) linear(a: step(-5))
251 #else
252 #pragma omp target simd private(argc, b), firstprivate(argv, c), lastprivate(d, f) collapse(2) if (target: argc) reduction(+ : h) linear(a:-5)
253 #endif
254 for (int i = 0; i < 10; ++i)
255 for (int j = 0; j < 10; ++j)
256 foo();
257 // CHECK: #pragma omp target simd private(argc,b) firstprivate(argv,c) lastprivate(d,f) collapse(2) if(target: argc) reduction(+: h) linear(a: step(-5))
258 // CHECK-NEXT: for (int i = 0; i < 10; ++i)
259 // CHECK-NEXT: for (int j = 0; j < 10; ++j)
260 // CHECK-NEXT: foo();
261
262 #pragma omp target simd private(argc,b) firstprivate(argv) if (argc > 0) reduction(+:c, arr1[argc]) reduction(max:e, arr[:5][0:10])
263 for (int i = 0; i < 2; ++i) {}
264 // CHECK: #pragma omp target simd private(argc,b) firstprivate(argv) if(argc > 0) reduction(+: c,arr1[argc]) reduction(max: e,arr[:5][0:10])
265 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
266 // CHECK-NEXT: }
267
268 #pragma omp target simd if (5) reduction(^:e, f, arr[0:5][:argc]) reduction(&& : h)
269 for (int i = 0; i < 2; ++i) {}
270 // CHECK: #pragma omp target simd if(5) reduction(^: e,f,arr[0:5][:argc]) reduction(&&: h)
271 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
272 // CHECK-NEXT: }
273
274 #if defined(OMP51) || defined(OMP52)
275 #pragma omp target simd if (target:argc > 0) if(simd:argc) nontemporal(argc, c, d) lastprivate(conditional: d, f) order(reproducible:concurrent)
276 #elif OMP5
277 #pragma omp target simd if (target:argc > 0) if(simd:argc) nontemporal(argc, c, d) lastprivate(conditional: d, f) order(concurrent)
278 #else
279 #pragma omp target simd if (target:argc > 0)
280 #endif // OMP51
281 for (int i = 0; i < 2; ++i) {}
282 // OMP45: #pragma omp target simd if(target: argc > 0)
283 // OMP50: #pragma omp target simd if(target: argc > 0) if(simd: argc) nontemporal(argc,c,d) lastprivate(conditional: d,f) order(concurrent)
284 // OMP51: #pragma omp target simd if(target: argc > 0) if(simd: argc) nontemporal(argc,c,d) lastprivate(conditional: d,f) order(reproducible: concurrent)
285 // OMP52: #pragma omp target simd if(target: argc > 0) if(simd: argc) nontemporal(argc,c,d) lastprivate(conditional: d,f) order(reproducible: concurrent)
286 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
287 // CHECK-NEXT: }
288
289 #pragma omp target simd if (5)
290 for (int i = 0; i < 2; ++i) {}
291 // CHECK: #pragma omp target simd if(5)
292 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
293 // CHECK-NEXT: }
294
295 #pragma omp target simd map(i)
296 for (int i = 0; i < 2; ++i) {}
297 // CHECK: #pragma omp target simd map(tofrom: i)
298 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
299 // CHECK-NEXT: }
300
301 #pragma omp target simd map(arr1[0:10], i)
302 for (int i = 0; i < 2; ++i) {}
303 // CHECK: #pragma omp target simd map(tofrom: arr1[0:10],i)
304 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
305 // CHECK-NEXT: }
306
307 #pragma omp target simd map(to: i) map(from: j)
308 for (int i = 0; i < 2; ++i) {}
309 // CHECK: #pragma omp target simd map(to: i) map(from: j)
310 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
311 // CHECK-NEXT: }
312
313 #pragma omp target simd map(always,alloc: i)
314 for (int i = 0; i < 2; ++i) {}
315 // CHECK: #pragma omp target simd map(always,alloc: i)
316 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
317 // CHECK-NEXT: }
318
319 #pragma omp target simd nowait
320 for (int i = 0; i < 2; ++i) {}
321 // CHECK: #pragma omp target simd nowait
322 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
323 // CHECK-NEXT: }
324
325 #pragma omp target simd depend(in : argc, arr[i:argc], arr1[:])
326 for (int i = 0; i < 2; ++i) {}
327 // CHECK: #pragma omp target simd depend(in : argc,arr[i:argc],arr1[:])
328 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
329 // CHECK-NEXT: }
330
331 #pragma omp target simd defaultmap(tofrom: scalar)
332 for (int i = 0; i < 2; ++i) {}
333 // CHECK: #pragma omp target simd defaultmap(tofrom: scalar)
334 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
335 // CHECK-NEXT: }
336
337 #pragma omp target simd safelen(clen-1)
338 for (int i = 0; i < 2; ++i) {}
339 // CHECK: #pragma omp target simd safelen(clen - 1)
340 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
341 // CHECK-NEXT: }
342
343 #pragma omp target simd simdlen(clen-1)
344 for (int i = 0; i < 2; ++i) {}
345 // CHECK: #pragma omp target simd simdlen(clen - 1)
346 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
347 // CHECK-NEXT: }
348
349 #pragma omp target simd aligned(arr1:4)
350 for (int i = 0; i < 2; ++i) {}
351 // CHECK: #pragma omp target simd aligned(arr1: 4)
352 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
353 // CHECK-NEXT: }
354
355 #pragma omp target simd is_device_ptr(p)
356 for (int i = 0; i < 2; ++i) {}
357 // CHECK: #pragma omp target simd is_device_ptr(p)
358 // CHECK-NEXT: for (int i = 0; i < 2; ++i) {
359 // CHECK-NEXT: }
360
361 return (tmain<int, 5>(argc, &argc));
362 }
363
364 #endif
365