1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
23
24
25
26
27
28
29
30
31
32
33
34
35
36
37
38
39
40
41
42
43
44
45
46
47
48
49
50
51
52
53
54
55
56
57
58
59
60
61
62
63
64
65
66
67
68
69
70
71
72
73
74
75
76
77
78
79
80
81
82
83
84
85
86
87
88
89
90
91
92
93
94
95
96
97
98
99
100
101
102
103
104
105
106
107
108
109
110
111
112
113
114
115
116
117
118
119
120
121
122
123
124
125
126
127
128
129
130
131
132
133
134
135
136
137
138
139
140
141
142
143
144
145
146
147
148
149
150
151
152
153
154
155
156
157
158
159
160
161
162
163
164
165
166
167
168
169
170
171
172
173
174
175
176
177
178
179
180
181
182
183
184
185
186
187
188
189
190
191
192
193
194
195
196
197
198
199
200
201
202
203
204
205
206
207
208
209
210
211
212
213
214
215
216
217
218
219
220
221
222
223
224
225
226
227
228
229
230
231
232
|
// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -fexceptions -fcxx-exceptions -o - | FileCheck %s
// RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple x86_64-unknown-unknown -fexceptions -fcxx-exceptions -emit-pch -o %t %s
// RUN: %clang_cc1 -fopenmp -x c++ -triple x86_64-unknown-unknown -fexceptions -fcxx-exceptions -debug-info-kind=limited -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s
// RUN: %clang_cc1 -verify -triple x86_64-apple-darwin10 -fopenmp -fexceptions -fcxx-exceptions -debug-info-kind=line-tables-only -x c++ -emit-llvm %s -o - | FileCheck %s --check-prefix=TERM_DEBUG
// expected-no-diagnostics
#ifndef HEADER
#define HEADER
long long get_val() { return 0; }
double *g_ptr;
// CHECK-LABEL: define {{.*void}} @{{.*}}simple{{.*}}(float* {{.+}}, float* {{.+}}, float* {{.+}}, float* {{.+}})
void simple(float *a, float *b, float *c, float *d) {
// CHECK: store i32 3, i32* %
// CHECK: icmp slt i32 %{{.+}}, 32
// CHECK: fmul float
// CHECK: fmul float
// CHECK: add nsw i32 %{{.+}}, 5
#pragma omp target parallel for simd device((int)*a)
for (int i = 3; i < 32; i += 5) {
a[i] = b[i] * c[i] * d[i];
}
// CHECK: call i{{.+}} @{{.+}}get_val{{.+}}()
// CHECK: store i32 10, i32* %
// CHECK: icmp sgt i32 %{{.+}}, 1
// CHECK: fadd float %{{.+}}, 1.000000e+00
// CHECK: add nsw {{.+}} %{{.+}}, 3
// CHECK: add nsw i32 %{{.+}}, -1
long long k = get_val();
#pragma omp target parallel for simd linear(k : 3) schedule(dynamic)
for (int i = 10; i > 1; i--) {
a[k]++;
k = k + 3;
}
// CHECK: store i32 12, i32* %
// CHECK: store i{{.+}} 2000, i{{.+}}* %
// CHECK: icmp uge i{{.+}} %{{.+}}, 600
// CHECK: store double 0.000000e+00,
// CHECK: fadd float %{{.+}}, 1.000000e+00
// CHECK: sub i{{.+}} %{{.+}}, 400
int lin = 12;
#pragma omp target parallel for simd linear(lin : get_val()), linear(g_ptr)
for (unsigned long long it = 2000; it >= 600; it-=400) {
*g_ptr++ = 0.0;
a[it + lin]++;
}
// CHECK: store i{{.+}} 6, i{{.+}}* %
// CHECK: icmp sle i{{.+}} %{{.+}}, 20
// CHECK: sub nsw i{{.+}} %{{.+}}, -4
#pragma omp target parallel for simd
for (short it = 6; it <= 20; it-=-4) {
}
// CHECK: store i8 122, i8* %
// CHECK: icmp sge i32 %{{.+}}, 97
// CHECK: add nsw i32 %{{.+}}, -1
#pragma omp target parallel for simd
for (unsigned char it = 'z'; it >= 'a'; it+=-1) {
}
// CHECK: store i32 100, i32* %
// CHECK: icmp ult i32 %{{.+}}, 10
// CHECK: add i32 %{{.+}}, 10
#pragma omp target parallel for simd
for (unsigned i=100; i<10; i+=10) {
}
int A;
{
A = -1;
// CHECK: store i{{.+}} -10, i{{.+}}* %
// CHECK: icmp slt i{{.+}} %{{.+}}, 10
// CHECK: add nsw i{{.+}} %{{.+}}, 3
#pragma omp target parallel for simd lastprivate(A)
for (long long i = -10; i < 10; i += 3) {
A = i;
}
}
int R;
{
R = -1;
// CHECK: store i{{.+}} -10, i{{.+}}* %
// CHECK: icmp slt i{{.+}} %{{.+}}, 10
// CHECK: add nsw i{{.+}} %{{.+}}, 3
#pragma omp target parallel for simd reduction(*:R)
for (long long i = -10; i < 10; i += 3) {
R *= i;
}
}
}
template <class T, unsigned K> T tfoo(T a) { return a + K; }
template <typename T, unsigned N>
int templ1(T a, T *z) {
#pragma omp target parallel for simd collapse(N)
for (int i = 0; i < N * 2; i++) {
for (long long j = 0; j < (N + N + N + N); j += 2) {
z[i + j] = a + tfoo<T, N>(i + j);
}
}
return 0;
}
// Instatiation templ1<float,2>
// CHECK-LABEL: define {{.*i32}} @{{.*}}templ1{{.*}}(float {{.+}}, float* {{.+}})
void inst_templ1() {
float a;
float z[100];
templ1<float,2> (a, z);
}
typedef int MyIdx;
class IterDouble {
double *Ptr;
public:
IterDouble operator++ () const {
IterDouble n;
n.Ptr = Ptr + 1;
return n;
}
bool operator < (const IterDouble &that) const {
return Ptr < that.Ptr;
}
double & operator *() const {
return *Ptr;
}
MyIdx operator - (const IterDouble &that) const {
return (MyIdx) (Ptr - that.Ptr);
}
IterDouble operator + (int Delta) {
IterDouble re;
re.Ptr = Ptr + Delta;
return re;
}
///~IterDouble() {}
};
// CHECK-LABEL: define {{.*void}} @{{.*}}iter_simple{{.*}}
void iter_simple(IterDouble ia, IterDouble ib, IterDouble ic) {
//
// Calculate number of iterations before the loop body.
// CHECK: invoke {{.*}}i32 @{{.*}}IterDouble{{.*}}
#pragma omp target parallel for simd
for (IterDouble i = ia; i < ib; ++i) {
// Call of operator+ (i, IV).
// CHECK: {{%.+}} = invoke {{.+}} @{{.*}}IterDouble{{.*}}
// ... loop body ...
*i = *ic * 0.5;
// Float multiply and save result.
// CHECK: [[MULR:%.+]] = fmul double {{%.+}}, 5.000000e-01
// CHECK-NEXT: invoke {{.+}} @{{.*}}IterDouble{{.*}}
// CHECK: store double [[MULR:%.+]], double* [[RESULT_ADDR:%.+]]
++ic;
}
// CHECK: ret void
}
// CHECK-LABEL: define {{.*void}} @{{.*}}collapsed{{.*}}
void collapsed(float *a, float *b, float *c, float *d) {
int i; // outer loop counter
unsigned j; // middle loop couter, leads to unsigned icmp in loop header.
// k declared in the loop init below
short l; // inner loop counter
//
#pragma omp target parallel for simd collapse(4)
for (i = 1; i < 3; i++) // 2 iterations
for (j = 2u; j < 5u; j++) //3 iterations
for (int k = 3; k <= 6; k++) // 4 iterations
for (l = 4; l < 9; ++l) // 5 iterations
{
// ... loop body ...
// End of body: store into a[i]:
// CHECK: store float [[RESULT:%.+]], float* [[RESULT_ADDR:%.+]]
float res = b[j] * c[k];
a[i] = res * d[l];
}
// CHECK: ret void
}
extern char foo();
extern double globalfloat;
// CHECK-LABEL: define {{.*void}} @{{.*}}widened{{.*}}
void widened(float *a, float *b, float *c, float *d) {
int i; // outer loop counter
short j; // inner loop counter
globalfloat = 1.0;
int localint = 1;
// CHECK: store double {{.+}}, double* [[GLOBALFLOAT:@.+]],
#pragma omp target parallel for simd collapse(2) private(globalfloat, localint)
for (i = 1; i < 3; i++) // 2 iterations
for (j = 0; j < foo(); j++) // foo() iterations
{
// ... loop body ...
//
// Here we expect store into private double var, not global
// CHECK: store double {{.+}}, double* [[GLOBALFLOAT]]
globalfloat = (float)j/i;
float res = b[j] * c[j];
// Store into a[i]:
// CHECK: store float [[RESULT:%.+]], float* [[RESULT_ADDR:%.+]]
a[i] = res * d[i];
// Then there's a store into private var localint:
// CHECK: store i32 {{.+}}, i32* [[LOCALINT:%[^,]+]]
localint = (int)j;
}
//
// Here we expect store into original localint, not its privatized version.
// CHECK: store i32 {{.+}}, i32* [[LOCALINT]]
localint = (int)j;
// CHECK: ret void
}
// TERM_DEBUG-LABEL: bar
int bar() {return 0;};
// TERM_DEBUG-LABEL: parallel_simd
void parallel_simd(float *a) {
#pragma omp target parallel for simd
for (unsigned i = 131071; i <= 2147483647; i += 127)
a[i] += bar();
}
#endif // HEADER
|
