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 -fsyntax-only -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 -fsyntax-only -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP50
7 
8 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -ast-print %s -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP45
9 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping
10 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP45
11 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=50 -DOMP5 -ast-print %s -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP50
12 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -DOMP5 -x c++ -std=c++11 -emit-pch -o %t %s -Wno-openmp-mapping
13 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=50 -DOMP5 -std=c++11 -include-pch %t -fsyntax-only -verify %s -ast-print -Wno-openmp-mapping | FileCheck %s --check-prefix=CHECK --check-prefix=OMP50
14 // expected-no-diagnostics
15 
16 #ifndef HEADER
17 #define HEADER
18 
19 typedef void **omp_allocator_handle_t;
20 extern const omp_allocator_handle_t omp_null_allocator;
21 extern const omp_allocator_handle_t omp_default_mem_alloc;
22 extern const omp_allocator_handle_t omp_large_cap_mem_alloc;
23 extern const omp_allocator_handle_t omp_const_mem_alloc;
24 extern const omp_allocator_handle_t omp_high_bw_mem_alloc;
25 extern const omp_allocator_handle_t omp_low_lat_mem_alloc;
26 extern const omp_allocator_handle_t omp_cgroup_mem_alloc;
27 extern const omp_allocator_handle_t omp_pteam_mem_alloc;
28 extern const omp_allocator_handle_t omp_thread_mem_alloc;
29 
foo()30 void foo() {}
31 
32 struct S {
SS33   S(): a(0) {}
SS34   S(int v) : a(v) {}
35   int a;
36   typedef int type;
37 };
38 
39 template <typename T>
40 class S7 : public T {
41 protected:
42   T a;
S7()43   S7() : a(0) {}
44 
45 public:
S7(typename T::type v)46   S7(typename T::type v) : a(v) {
47 #pragma omp target teams distribute simd private(a) private(this->a) private(T::a)
48     for (int k = 0; k < a.a; ++k)
49       ++this->a.a;
50   }
operator =(S7 & s)51   S7 &operator=(S7 &s) {
52     int k;
53 #pragma omp target teams distribute simd allocate(a) private(a) private(this->a) linear(k) allocate(k)
54     for (k = 0; k < s.a.a; ++k)
55       ++s.a.a;
56     return *this;
57   }
foo()58   void foo() {
59     int b, argv, d, c, e, f;
60 #pragma omp target teams distribute simd private(b), firstprivate(argv) shared(d) reduction(+:c) reduction(max:e) num_teams(f) thread_limit(d)
61     for (int k = 0; k < a.a; ++k)
62       ++a.a;
63   }
bar()64   void bar() {
65     int arr[10];
66     const int alen = 16;
67     const int slen1 = 8;
68     const int slen2 = 8;
69 #pragma omp target teams distribute simd simdlen(slen1) safelen(slen2) aligned(arr:alen)
70     for (int k = 0; k < a.a; ++k)
71       ++a.a;
72   }
73 };
74 // CHECK: #pragma omp target teams distribute simd private(this->a) private(this->a) private(T::a)
75 // CHECK: #pragma omp target teams distribute simd allocate(this->a) private(this->a) private(this->a) linear(k) allocate(k)
76 // CHECK: #pragma omp target teams distribute simd private(b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d)
77 // CHECK: #pragma omp target teams distribute simd simdlen(slen1) safelen(slen2) aligned(arr: alen)
78 // CHECK: #pragma omp target teams distribute simd private(this->a) private(this->a) private(this->S::a)
79 
80 class S8 : public S7<S> {
S8()81   S8() {}
82 
83 public:
S8(int v)84   S8(int v) : S7<S>(v){
85 #pragma omp target teams distribute simd private(a) private(this->a) private(S7<S>::a)
86     for (int k = 0; k < a.a; ++k)
87       ++this->a.a;
88   }
operator =(S8 & s)89   S8 &operator=(S8 &s) {
90 #pragma omp target teams distribute simd private(a) private(this->a)
91     for (int k = 0; k < s.a.a; ++k)
92       ++s.a.a;
93     return *this;
94   }
bar()95   void bar() {
96     int b, argv, d, c, e, f;
97 #pragma omp target teams distribute simd private(b), firstprivate(argv) shared(d) reduction(+:c) reduction(max:e) num_teams(f) thread_limit(d)
98     for (int k = 0; k < a.a; ++k)
99       ++a.a;
100   }
foo()101   void foo() {
102     const int alen = 16;
103     const int slen1 = 8;
104     const int slen2 = 8;
105     int arr[10];
106 #pragma omp target teams distribute simd simdlen(slen1) safelen(slen2) aligned(arr:alen)
107     for (int k = 0; k < a.a; ++k)
108       ++a.a;
109   }
110 };
111 // CHECK: #pragma omp target teams distribute simd private(this->a) private(this->a) private(this->S7<S>::a)
112 // CHECK: #pragma omp target teams distribute simd private(this->a) private(this->a)
113 // CHECK: #pragma omp target teams distribute simd private(b) firstprivate(argv) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d)
114 // CHECK: #pragma omp target teams distribute simd simdlen(slen1) safelen(slen2) aligned(arr: alen)
115 
116 template <class T, int N>
117 T tmain(T argc) {
118   T b = argc, c, d, e, f, g;
119   static T a;
120 // CHECK: static T a;
121   const T clen = 5;
122   const T alen = 16;
123   int arr[10];
124 #pragma omp target teams distribute simd
125   for (int i=0; i < 2; ++i)
126     a = 2;
127 // CHECK: #pragma omp target teams distribute simd{{$}}
128 // CHECK-NEXT: for (int i = 0; i < 2; ++i)
129 // CHECK-NEXT: a = 2;
130 #pragma omp target teams distribute simd private(argc, b), firstprivate(c, d), collapse(2)
131   for (int i = 0; i < 10; ++i)
132     for (int j = 0; j < 10; ++j)
133       foo();
134 // CHECK: #pragma omp target teams distribute simd private(argc,b) firstprivate(c,d) collapse(2)
135 // CHECK-NEXT: for (int i = 0; i < 10; ++i)
136 // CHECK-NEXT: for (int j = 0; j < 10; ++j)
137 // CHECK-NEXT: foo();
138   for (int i = 0; i < 10; ++i)
139     foo();
140 // CHECK: for (int i = 0; i < 10; ++i)
141 // CHECK-NEXT: foo();
142 #pragma omp target teams distribute simd
143   for (int i = 0; i < 10; ++i)
144     foo();
145 // CHECK: #pragma omp target teams distribute simd
146 // CHECK-NEXT: for (int i = 0; i < 10; ++i)
147 // CHECK-NEXT: foo();
148 #pragma omp target teams distribute simd private(b), firstprivate(argc) shared(d) reduction(+:c) reduction(max:e) num_teams(f) thread_limit(d)
149     for (int k = 0; k < 10; ++k)
150       e += d + argc;
151 // CHECK: #pragma omp target teams distribute simd private(b) firstprivate(argc) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d)
152 // CHECK-NEXT: for (int k = 0; k < 10; ++k)
153 // CHECK-NEXT: e += d + argc;
154 #pragma omp target teams distribute simd simdlen(clen-1)
155   for (int k = 0; k < 10; ++k)
156     e += d + argc;
157 // CHECK: #pragma omp target teams distribute simd simdlen(clen - 1)
158 // CHECK-NEXT: for (int k = 0; k < 10; ++k)
159 // CHECK-NEXT: e += d + argc;
160 #pragma omp target teams distribute simd safelen(clen-1) aligned(arr:alen)
161   for (int k = 0; k < 10; ++k)
162     e += d + argc + arr[k];
163 // CHECK: #pragma omp target teams distribute simd safelen(clen - 1) aligned(arr: alen)
164 // CHECK-NEXT: for (int k = 0; k < 10; ++k)
165 // CHECK-NEXT: e += d + argc + arr[k];
166   return T();
167 }
168 
main(int argc,char ** argv)169 int main (int argc, char **argv) {
170   int b = argc, c, d, e, f, g;
171   static int a;
172 // CHECK: static int a;
173   const int clen = 5;
174   const int N = 10;
175   int arr[10];
176 #pragma omp target teams distribute simd
177   for (int i=0; i < 2; ++i)
178     a = 2;
179 // CHECK: #pragma omp target teams distribute simd
180 // CHECK-NEXT: for (int i = 0; i < 2; ++i)
181 // CHECK-NEXT: a = 2;
182 #pragma omp target teams distribute simd private(argc,b),firstprivate(argv, c), collapse(2)
183   for (int i = 0; i < 10; ++i)
184     for (int j = 0; j < 10; ++j)
185       foo();
186 // CHECK: #pragma omp target teams distribute simd private(argc,b) firstprivate(argv,c) collapse(2)
187 // CHECK-NEXT: for (int i = 0; i < 10; ++i)
188 // CHECK-NEXT: for (int j = 0; j < 10; ++j)
189 // CHECK-NEXT: foo();
190   for (int i = 0; i < 10; ++i)
191     foo();
192 // CHECK: for (int i = 0; i < 10; ++i)
193 // CHECK-NEXT: foo();
194 #pragma omp target teams distribute simd
195   for (int i = 0; i < 10; ++i)foo();
196 // CHECK: #pragma omp target teams distribute simd
197 // CHECK-NEXT: for (int i = 0; i < 10; ++i)
198 // CHECK-NEXT: foo();
199 #pragma omp target teams distribute simd private(b), firstprivate(argc) shared(d) reduction(+:c) reduction(max:e) num_teams(f) thread_limit(d)
200   for (int k = 0; k < 10; ++k)
201     e += d + argc;
202 // CHECK: #pragma omp target teams distribute simd private(b) firstprivate(argc) shared(d) reduction(+: c) reduction(max: e) num_teams(f) thread_limit(d)
203 // CHECK-NEXT: for (int k = 0; k < 10; ++k)
204 // CHECK-NEXT: e += d + argc;
205 #pragma omp target teams distribute simd simdlen(clen-1)
206   for (int k = 0; k < 10; ++k)
207     e += d + argc;
208 // CHECK: #pragma omp target teams distribute simd simdlen(clen - 1)
209 // CHECK-NEXT: for (int k = 0; k < 10; ++k)
210 // CHECK-NEXT: e += d + argc;
211 #ifdef OMP5
212 #pragma omp target teams distribute simd safelen(clen-1) aligned(arr:N+6) if(simd:argc) nontemporal(argc, c, d) order(concurrent) allocate(omp_low_lat_mem_alloc:e) firstprivate(e) uses_allocators(omp_low_lat_mem_alloc)
213 #else
214 #pragma omp target teams distribute simd safelen(clen-1) aligned(arr:N+6)
215 #endif // OMP5
216   for (int k = 0; k < 10; ++k)
217     e += d + argc + arr[k];
218 // OMP45: #pragma omp target teams distribute simd safelen(clen - 1) aligned(arr: N + 6)
219 // OMP50: #pragma omp target teams distribute simd safelen(clen - 1) aligned(arr: N + 6) if(simd: argc) nontemporal(argc,c,d) order(concurrent) allocate(omp_low_lat_mem_alloc: e) firstprivate(e) uses_allocators(omp_low_lat_mem_alloc)
220 // CHECK-NEXT: for (int k = 0; k < 10; ++k)
221 // CHECK-NEXT: e += d + argc + arr[k];
222   return (0);
223 }
224 
225 #endif
226