1*f82ec553SMike Rice // RUN: %clang_cc1 -triple x86_64-pc-linux-gnu -fopenmp -fopenmp-version=51 \
2*f82ec553SMike Rice // RUN:   -fsyntax-only -verify %s
3*f82ec553SMike Rice 
4*f82ec553SMike Rice // RUN: %clang_cc1 -triple x86_64-pc-linux-gnu -fopenmp -fopenmp-version=51 \
5*f82ec553SMike Rice // RUN:   -ast-print %s | FileCheck %s
6*f82ec553SMike Rice 
7*f82ec553SMike Rice // RUN: %clang_cc1 -triple x86_64-pc-linux-gnu -fopenmp -fopenmp-version=51 \
8*f82ec553SMike Rice // RUN:   -emit-pch -o %t %s
9*f82ec553SMike Rice 
10*f82ec553SMike Rice // RUN: %clang_cc1 -triple x86_64-pc-linux-gnu -fopenmp -fopenmp-version=51 \
11*f82ec553SMike Rice // RUN:   -include-pch %t -ast-print %s | FileCheck %s
12*f82ec553SMike Rice 
13*f82ec553SMike Rice // expected-no-diagnostics
14*f82ec553SMike Rice 
15*f82ec553SMike Rice #ifndef HEADER
16*f82ec553SMike Rice #define HEADER
17*f82ec553SMike Rice 
18*f82ec553SMike Rice typedef void **omp_allocator_handle_t;
19*f82ec553SMike Rice extern const omp_allocator_handle_t omp_null_allocator;
20*f82ec553SMike Rice extern const omp_allocator_handle_t omp_default_mem_alloc;
21*f82ec553SMike Rice extern const omp_allocator_handle_t omp_large_cap_mem_alloc;
22*f82ec553SMike Rice extern const omp_allocator_handle_t omp_const_mem_alloc;
23*f82ec553SMike Rice extern const omp_allocator_handle_t omp_high_bw_mem_alloc;
24*f82ec553SMike Rice extern const omp_allocator_handle_t omp_low_lat_mem_alloc;
25*f82ec553SMike Rice extern const omp_allocator_handle_t omp_cgroup_mem_alloc;
26*f82ec553SMike Rice extern const omp_allocator_handle_t omp_pteam_mem_alloc;
27*f82ec553SMike Rice extern const omp_allocator_handle_t omp_thread_mem_alloc;
28*f82ec553SMike Rice 
29*f82ec553SMike Rice //CHECK: template <typename T, int C, int D> void templ_foo(T t) {
30*f82ec553SMike Rice //CHECK:   T j, z;
31*f82ec553SMike Rice //CHECK:   #pragma omp target parallel loop device(D) collapse(C) reduction(+: z) lastprivate(j) bind(thread) num_threads(C + 2)
32*f82ec553SMike Rice //CHECK:   for (T i = 0; i < t; ++i)
33*f82ec553SMike Rice //CHECK:       for (j = 0; j < t; ++j)
34*f82ec553SMike Rice //CHECK:           z += i + j;
35*f82ec553SMike Rice //CHECK: }
36*f82ec553SMike Rice 
37*f82ec553SMike Rice //CHECK: template<> void templ_foo<int, 2, 0>(int t) {
38*f82ec553SMike Rice //CHECK:     int j, z;
39*f82ec553SMike Rice //CHECK:     #pragma omp target parallel loop device(0) collapse(2) reduction(+: z) lastprivate(j) bind(thread) num_threads(2 + 2)
40*f82ec553SMike Rice //CHECK:         for (int i = 0; i < t; ++i)
41*f82ec553SMike Rice //CHECK:             for (j = 0; j < t; ++j)
42*f82ec553SMike Rice //CHECK:                 z += i + j;
43*f82ec553SMike Rice //CHECK: }
44*f82ec553SMike Rice template <typename T, int C, int D>
templ_foo(T t)45*f82ec553SMike Rice void templ_foo(T t) {
46*f82ec553SMike Rice 
47*f82ec553SMike Rice   T j,z;
48*f82ec553SMike Rice   #pragma omp target parallel loop device(D) collapse(C) reduction(+:z) lastprivate(j) bind(thread) num_threads(C+2)
49*f82ec553SMike Rice   for (T i = 0; i<t; ++i)
50*f82ec553SMike Rice     for (j = 0; j<t; ++j)
51*f82ec553SMike Rice       z += i+j;
52*f82ec553SMike Rice }
53*f82ec553SMike Rice 
54*f82ec553SMike Rice 
55*f82ec553SMike Rice //CHECK: void test() {
test()56*f82ec553SMike Rice void test() {
57*f82ec553SMike Rice   constexpr int N = 100;
58*f82ec553SMike Rice   float MTX[N][N];
59*f82ec553SMike Rice   int aaa[1000];
60*f82ec553SMike Rice 
61*f82ec553SMike Rice   //CHECK: #pragma omp target parallel loop map(tofrom: MTX)
62*f82ec553SMike Rice   #pragma omp target parallel loop map(MTX)
63*f82ec553SMike Rice   for (auto j = 0; j < N; ++j) {
64*f82ec553SMike Rice     MTX[0][j] = 0;
65*f82ec553SMike Rice   }
66*f82ec553SMike Rice 
67*f82ec553SMike Rice   int j, z, z1;
68*f82ec553SMike Rice   //CHECK: #pragma omp target parallel loop collapse(2) private(z) lastprivate(j) order(concurrent) reduction(+: z1) bind(parallel)
69*f82ec553SMike Rice   #pragma omp target parallel loop collapse(2) private(z) lastprivate(j) \
70*f82ec553SMike Rice                          order(concurrent) reduction(+:z1) bind(parallel)
71*f82ec553SMike Rice   for (auto i = 0; i < N; ++i) {
72*f82ec553SMike Rice     for (j = 0; j < N; ++j) {
73*f82ec553SMike Rice       z = i+j;
74*f82ec553SMike Rice       MTX[i][j] = z;
75*f82ec553SMike Rice       z1 += z;
76*f82ec553SMike Rice     }
77*f82ec553SMike Rice   }
78*f82ec553SMike Rice 
79*f82ec553SMike Rice   //CHECK: #pragma omp target parallel loop bind(thread) num_threads(16) default(none)
80*f82ec553SMike Rice   #pragma omp target parallel loop bind(thread) num_threads(16) default(none)
81*f82ec553SMike Rice   for (auto i = 0; i < N; ++i) { }
82*f82ec553SMike Rice 
83*f82ec553SMike Rice   int pr;
84*f82ec553SMike Rice   int zzz;
85*f82ec553SMike Rice   //CHECK: #pragma omp target parallel loop private(zzz) uses_allocators(omp_default_mem_alloc) allocate(omp_default_mem_alloc: zzz) if(1) device(0) proc_bind(close) map(tofrom: pr)
86*f82ec553SMike Rice   #pragma omp target parallel loop private(zzz) uses_allocators(omp_default_mem_alloc) allocate(omp_default_mem_alloc:zzz) if(1) device(0) proc_bind(close) map(tofrom:pr)
87*f82ec553SMike Rice   for (int i=0; i<1000; ++i) {
88*f82ec553SMike Rice     zzz = i + 1;
89*f82ec553SMike Rice     pr = 33;
90*f82ec553SMike Rice   }
91*f82ec553SMike Rice 
92*f82ec553SMike Rice   int fpr = 10;
93*f82ec553SMike Rice   int k;
94*f82ec553SMike Rice   int s = 20;
95*f82ec553SMike Rice   //CHECK: #pragma omp target parallel loop bind(thread) private(pr) firstprivate(fpr) shared(s) allocate(k) reduction(+: k)
96*f82ec553SMike Rice   #pragma omp target parallel loop bind(thread) private(pr) firstprivate(fpr) \
97*f82ec553SMike Rice                         shared(s) allocate(k)  reduction(+:k)
98*f82ec553SMike Rice   for (auto i = 0; i < N; ++i) {
99*f82ec553SMike Rice     pr = i + fpr + s;
100*f82ec553SMike Rice   }
101*f82ec553SMike Rice 
102*f82ec553SMike Rice   short y = 3;
103*f82ec553SMike Rice   //CHECK: #pragma omp target parallel loop map(tofrom: y) depend(out : y)
104*f82ec553SMike Rice   #pragma omp target parallel loop map(tofrom:y) depend(out:y)
105*f82ec553SMike Rice   for (int i=0; i<10; ++i) {
106*f82ec553SMike Rice     y = 3+i;
107*f82ec553SMike Rice   }
108*f82ec553SMike Rice }
109*f82ec553SMike Rice 
110*f82ec553SMike Rice //CHECK: void nobindingfunc() {
nobindingfunc()111*f82ec553SMike Rice void nobindingfunc()
112*f82ec553SMike Rice {
113*f82ec553SMike Rice   //CHECK: #pragma omp target parallel loop
114*f82ec553SMike Rice   #pragma omp target parallel loop
115*f82ec553SMike Rice   for (int i=0; i<10; ++i) { }
116*f82ec553SMike Rice }
117*f82ec553SMike Rice 
bar()118*f82ec553SMike Rice void bar()
119*f82ec553SMike Rice {
120*f82ec553SMike Rice   templ_foo<int,2,0>(8);
121*f82ec553SMike Rice }
122*f82ec553SMike Rice 
123*f82ec553SMike Rice #endif // HEADER
124