1 | // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck %s --check-prefix LAMBDA --check-prefix LAMBDA-64 |
2 | // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s |
3 | // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck %s --check-prefix LAMBDA --check-prefix LAMBDA-64 |
4 | // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck %s --check-prefix LAMBDA --check-prefix LAMBDA-32 |
5 | // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s |
6 | // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck %s --check-prefix LAMBDA --check-prefix LAMBDA-32 |
7 | |
8 | // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY0 %s |
9 | // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s |
10 | // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY0 %s |
11 | // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY0 %s |
12 | // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s |
13 | // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY0 %s |
14 | // SIMD-ONLY0-NOT: {{__kmpc|__tgt}} |
15 | |
16 | // RUN: %clang_cc1 -verify -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64 |
17 | // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s |
18 | // RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64 |
19 | // RUN: %clang_cc1 -verify -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32 |
20 | // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s |
21 | // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32 |
22 | |
23 | // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY1 %s |
24 | // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s |
25 | // RUN: %clang_cc1 -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY1 %s |
26 | // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY1 %s |
27 | // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s |
28 | // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - -Wno-openmp-target | FileCheck --check-prefix SIMD-ONLY1 %s |
29 | // SIMD-ONLY1-NOT: {{__kmpc|__tgt}} |
30 | // expected-no-diagnostics |
31 | #ifndef HEADER |
32 | #define HEADER |
33 | |
34 | template <class T> |
35 | struct S { |
36 | T f; |
37 | S(T a) : f(a) {} |
38 | S() : f() {} |
39 | operator T() { return T(); } |
40 | ~S() {} |
41 | }; |
42 | |
43 | // CHECK: [[S_FLOAT_TY:%.+]] = type { float } |
44 | // CHECK: [[S_INT_TY:%.+]] = type { i{{[0-9]+}} } |
45 | template <typename T> |
46 | T tmain() { |
47 | S<T> test; |
48 | T t_var = T(); |
49 | T vec[] = {1, 2}; |
50 | S<T> s_arr[] = {1, 2}; |
51 | S<T> &var = test; |
52 | #pragma omp target |
53 | #pragma omp teams |
54 | #pragma omp distribute simd firstprivate(t_var, vec, s_arr, s_arr, var, var) |
55 | for (int i = 0; i < 2; ++i) { |
56 | vec[i] = t_var; |
57 | s_arr[i] = var; |
58 | } |
59 | return T(); |
60 | } |
61 | |
62 | int main() { |
63 | static int svar; |
64 | volatile double g; |
65 | volatile double &g1 = g; |
66 | |
67 | #ifdef LAMBDA |
68 | // LAMBDA-LABEL: @main |
69 | // LAMBDA: call{{.*}} void [[OUTER_LAMBDA:@.+]]( |
70 | [&]() { |
71 | static float sfvar; |
72 | // LAMBDA: define{{.*}} internal{{.*}} void [[OUTER_LAMBDA]]( |
73 | // LAMBDA: call i{{[0-9]+}} @__tgt_target_teams( |
74 | // LAMBDA: call void [[OFFLOADING_FUN:@.+]]( |
75 | |
76 | // LAMBDA: define{{.+}} void [[OFFLOADING_FUN]]( |
77 | // LAMBDA: call {{.*}}void {{.+}} @__kmpc_fork_teams({{.+}}, {{.+}}, {{.+}}* [[OMP_OUTLINED:@.+]] to {{.+}}) |
78 | #pragma omp target |
79 | #pragma omp teams |
80 | #pragma omp distribute simd firstprivate(g, g1, svar, sfvar) |
81 | for (int i = 0; i < 2; ++i) { |
82 | // LAMBDA: define internal{{.*}} void [[OMP_OUTLINED]](i32* noalias %{{.+}}, i32* noalias %{{.+}}, double*{{.*}} [[G_IN:%.+]], double*{{.+}} [[G1_IN:%.+]], i{{[0-9]+}}*{{.+}} [[SVAR_IN:%.+]], float*{{.+}} [[SFVAR_IN:%.+]]) |
83 | // Private alloca's for conversion |
84 | // LAMBDA: [[G_ADDR:%.+]] = alloca double*, |
85 | // LAMBDA: [[G1_ADDR:%.+]] = alloca double*, |
86 | // LAMBDA: [[SVAR_ADDR:%.+]] = alloca i{{[0-9]+}}*, |
87 | // LAMBDA: [[SFVAR_ADDR:%.+]] = alloca float*, |
88 | // LAMBDA: [[G1_REF:%.+]] = alloca double*, |
89 | |
90 | // Actual private variables to be used in the body (tmp is used for the reference type) |
91 | // LAMBDA: [[G_PRIVATE:%.+]] = alloca double, |
92 | // LAMBDA: [[G1_PRIVATE:%.+]] = alloca double, |
93 | // LAMBDA: [[TMP_PRIVATE:%.+]] = alloca double*, |
94 | // LAMBDA: [[SVAR_PRIVATE:%.+]] = alloca i{{[0-9]+}}, |
95 | // LAMBDA: [[SFVAR_PRIVATE:%.+]] = alloca float, |
96 | |
97 | // Store input parameter addresses into private alloca's for conversion |
98 | // LAMBDA: store double* [[G_IN]], double** [[G_ADDR]], |
99 | // LAMBDA: store double* [[G1_IN]], double** [[G1_ADDR]], |
100 | // LAMBDA: store i{{[0-9]+}}* [[SVAR_IN]], i{{[0-9]+}}** [[SVAR_ADDR]], |
101 | // LAMBDA: store float* [[SFVAR_IN]], float** [[SFVAR_ADDR]], |
102 | |
103 | // LAMBDA-DAG: [[G_ADDR_VAL:%.+]] = load double*, double** [[G_ADDR]], |
104 | // LAMBDA-DAG: [[G1_ADDR_VAL:%.+]] = load double*, double** [[G1_ADDR]], |
105 | // LAMBDA-DAG: [[SVAR_ADDR_VAL:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[SVAR_ADDR]], |
106 | // LAMBDA-DAG: [[SFVAR_ADDR_VAL:%.+]] = load float*, float** [[SFVAR_ADDR]], |
107 | // LAMBDA-DAG: store double* [[G1_ADDR_VAL]], double** [[G1_REF]], |
108 | |
109 | // LAMBDA-DAG: [[G_VAL:%.+]] = load{{.+}} double, double* [[G_ADDR_VAL]], |
110 | // LAMBDA-DAG: store double [[G_VAL]], double* [[G_PRIVATE]], |
111 | // LAMBDA-DAG: [[G1_VAL_REF:%.+]] = load double*, double** [[G1_REF]], |
112 | // LAMBDA-DAG: [[G1_VAL:%.+]] = load{{.+}} double, double* [[G1_VAL_REF]], |
113 | // LAMBDA-DAG: store double [[G1_VAL]], double* [[G1_PRIVATE]], |
114 | // LAMBDA-DAG: store double* [[G1_PRIVATE]], double** [[TMP_PRIVATE]], |
115 | // LAMBDA-DAG: [[SVAR_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[SVAR_ADDR_VAL]], |
116 | // LAMBDA-DAG: store i{{[0-9]+}} [[SVAR_VAL]], i{{[0-9]+}}* [[SVAR_PRIVATE]], |
117 | // LAMBDA-DAG: [[SFVAR_VAL:%.+]] = load float, float* [[SFVAR_ADDR_VAL]], |
118 | // LAMBDA-DAG: store float [[SFVAR_VAL]], float* [[SFVAR_PRIVATE]], |
119 | |
120 | // LAMBDA: call {{.*}}void @__kmpc_for_static_init_4( |
121 | g += 1; |
122 | g1 += 1; |
123 | svar += 3; |
124 | sfvar += 4.0; |
125 | // LAMBDA-DAG: [[G_VAL:%.+]] = load double, double* [[G_PRIVATE]], |
126 | // LAMBDA-DAG: [[G_NEXT:%.+]] = fadd double [[G_VAL]], 1.{{.+}} |
127 | // LAMBDA-DAG: store double [[G_NEXT]], double* [[G_PRIVATE]], |
128 | // LAMBDA-DAG: [[TMP_VAL1:%.+]] = load double*, double** [[TMP_PRIVATE]], |
129 | // LAMBDA-DAG: [[TMP_VAL_VAL1:%.+]] = load{{.*}} double, double* [[TMP_VAL1]], |
130 | // LAMBDA-DAG: [[TMP_ADD:%.+]] = fadd double [[TMP_VAL_VAL1]], 1.{{.+}} |
131 | // LAMBDA-DAG: store{{.*}} double [[TMP_ADD]], double* [[TMP_VAL1]], |
132 | // LAMBDA-DAG: [[SVAR_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[SVAR_PRIVATE]], |
133 | // LAMBDA-DAG: [[SVAR_ADD:%.+]] = add{{.*}} i{{[0-9]+}} [[SVAR_VAL]], 3 |
134 | // LAMBDA-DAG: store i{{[0-9]+}} [[SVAR_ADD]], i{{[0-9]+}}* [[SVAR_PRIVATE]], |
135 | // LAMBDA-DAG: [[SFVAR_VAL:%.+]] = load float, float* [[SFVAR_PRIVATE]], |
136 | // LAMBDA-DAG: [[SFVAR_CONV_VAL1:%.+]] = fpext float [[SFVAR_VAL]] to double |
137 | // LAMBDA-DAG: [[SFVAR_ADD:%.+]] = fadd double [[SFVAR_CONV_VAL1]], 4.{{.+}} |
138 | // LAMBDA-DAG: [[SFVAR_CONV_VAL2:%.+]] = fptrunc double [[SFVAR_ADD]] to float |
139 | // LAMBDA-DAG: store float [[SFVAR_CONV_VAL2:%.+]], float* [[SFVAR_PRIVATE]], |
140 | |
141 | // call inner lambda (use refs to private alloca's) |
142 | // LAMBDA: [[GEP_0:%.+]] = getelementptr{{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 0 |
143 | // LAMBDA: store double* [[G_PRIVATE]], double** [[GEP_0]], |
144 | // LAMBDA: [[GEP_1:%.+]] = getelementptr{{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 1 |
145 | // LAMBDA: [[TMP_PAR:%.+]] = load double*, double** [[TMP_PRIVATE]], |
146 | // LAMBDA: store double* [[TMP_PAR]], double** [[GEP_1]], |
147 | // LAMBDA: [[GEP_2:%.+]] = getelementptr{{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 2 |
148 | // LAMBDA: store i{{[0-9]+}}* [[SVAR_PRIVATE]], i{{[0-9]+}}** [[GEP_2]], |
149 | // LAMBDA: [[GEP_3:%.+]] = getelementptr{{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 3 |
150 | // LAMBDA: store float* [[SFVAR_PRIVATE]], float** [[GEP_3]], |
151 | // LAMBDA: call{{.*}} void [[INNER_LAMBDA:@.+]](%{{.+}}* {{.+}}) |
152 | // LAMBDA: call {{.*}}void @__kmpc_for_static_fini( |
153 | [&]() { |
154 | // LAMBDA: define {{.+}} void [[INNER_LAMBDA]](%{{.+}}* [[ARG_PTR:%.+]]) |
155 | // LAMBDA: store %{{.+}}* [[ARG_PTR]], %{{.+}}** [[ARG_PTR_REF:%.+]], |
156 | g += 2; |
157 | g1 += 2; |
158 | svar += 4; |
159 | sfvar += 8.0; |
160 | // LAMBDA-DAG: [[ARG_PTR:%.+]] = load %{{.+}}*, %{{.+}}** [[ARG_PTR_REF]] |
161 | // LAMBDA-DAG: [[G_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 0 |
162 | // LAMBDA-DAG: [[G_REF:%.+]] = load double*, double** [[G_PTR_REF]], |
163 | // LAMBDA-DAG: [[G_REF_VAL:%.+]] = load double, double* [[G_REF]], |
164 | // LAMBDA-DAG: [[G_REF_ADD:%.+]] = fadd double [[G_REF_VAL]], 2.{{.+}} |
165 | // LAMBDA-DAG: store double [[G_REF_ADD]], double* [[G_REF]] |
166 | |
167 | // LAMBDA-DAG: [[TMP_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 1 |
168 | // LAMBDA-DAG: [[G1_REF:%.+]] = load double*, double** [[TMP_PTR_REF]] |
169 | // LAMBDA-DAG: [[G1_REF_VAL:%.+]] = load double, double* [[G1_REF]], |
170 | // LAMBDA-DAG: [[G1_ADD:%.+]] = fadd double [[G1_REF_VAL]], 2.{{.+}} |
171 | // LAMBDA-DAG: store double [[G1_ADD]], double* [[G1_REF]], |
172 | |
173 | // LAMBDA-DAG: [[SVAR_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 2 |
174 | // LAMBDA-DAG: [[SVAR_REF:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[SVAR_PTR_REF]] |
175 | // LAMBDA-DAG: [[SVAR_REF_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[SVAR_REF]] |
176 | // LAMBDA-DAG: [[SVAR_ADD:%.+]] = add{{.*}} i{{[0-9]+}} [[SVAR_REF_VAL]], 4 |
177 | // LAMBDA-DAG: store i{{[0-9]+}} [[SVAR_ADD]], i{{[0-9]+}}* [[SVAR_REF]] |
178 | |
179 | // LAMBDA-DAG: [[SFVAR_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 3 |
180 | // LAMBDA-DAG: [[SFVAR_REF:%.+]] = load float*, float** [[SFVAR_PTR_REF]] |
181 | // LAMBDA-DAG: [[SFVAR_REF_VAL:%.+]] = load float, float* [[SFVAR_REF]] |
182 | // LAMBDA-DAG: [[SFVAR_REF_CONV:%.+]] = fpext float [[SFVAR_REF_VAL]] to double |
183 | // LAMBDA-DAG: [[SFVAR_ADD:%.+]] = fadd double [[SFVAR_REF_CONV]], 8.{{.+}} |
184 | // LAMBDA-DAG: [[SFVAR_ADD_CONV:%.+]] = fptrunc double [[SFVAR_ADD]] to float |
185 | // LAMBDA-DAG: store float [[SFVAR_ADD_CONV]], float* [[SFVAR_REF]], |
186 | }(); |
187 | } |
188 | }(); |
189 | return 0; |
190 | #else |
191 | S<float> test; |
192 | int t_var = 0; |
193 | int vec[] = {1, 2}; |
194 | S<float> s_arr[] = {1, 2}; |
195 | S<float> &var = test; |
196 | |
197 | #pragma omp target |
198 | #pragma omp teams |
199 | #pragma omp distribute simd firstprivate(t_var, vec, s_arr, s_arr, var, var, svar) |
200 | for (int i = 0; i < 2; ++i) { |
201 | vec[i] = t_var; |
202 | s_arr[i] = var; |
203 | } |
204 | return tmain<int>(); |
205 | #endif |
206 | } |
207 | |
208 | // CHECK: define{{.*}} i{{[0-9]+}} @main() |
209 | // CHECK: [[TEST:%.+]] = alloca [[S_FLOAT_TY]], |
210 | // CHECK: call {{.*}} [[S_FLOAT_TY_DEF_CONSTR:@.+]]([[S_FLOAT_TY]]* [[TEST]]) |
211 | // CHECK: call i{{[0-9]+}} @__tgt_target_teams( |
212 | // CHECK: call void [[OFFLOAD_FUN:@.+]]( |
213 | // CHECK: ret |
214 | |
215 | // CHECK: define{{.+}} [[OFFLOAD_FUN]]( |
216 | // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 5, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*, i{{[0-9]+}}*, [2 x i{{[0-9]+}}]*, [2 x [[S_FLOAT_TY]]]*, [[S_FLOAT_TY]]*, i{{[0-9]+}}*)* [[OMP_OUTLINED:@.+]] to void |
217 | // CHECK: ret |
218 | // |
219 | // CHECK: define internal void [[OMP_OUTLINED]](i{{[0-9]+}}*{{.+}}, i{{[0-9]+}}*{{.+}}, i{{[0-9]+}}*{{.+}} [[T_VAR_IN:%.+]], [2 x i{{[0-9]+}}]*{{.*}} [[VEC_IN:%.+]], [2 x [[S_FLOAT_TY]]]*{{.*}} [[S_ARR_IN:%.+]], [[S_FLOAT_TY]]*{{.*}} [[VAR_IN:%.+]], i{{[0-9]+}}*{{.+}} [[SVAR_IN:%.+]]) |
220 | |
221 | // CHECK: alloca i{{[0-9]+}}*, |
222 | // CHECK: alloca i{{[0-9]+}}*, |
223 | // CHECK: [[T_VAR_ADDR:%.+]] = alloca i{{[0-9]+}}*, |
224 | // CHECK: [[VEC_ADDR:%.+]] = alloca [2 x i{{[0-9]+}}]*, |
225 | // CHECK: [[S_ARR_ADDR:%.+]] = alloca [2 x [[S_FLOAT_TY]]]*, |
226 | // CHECK: [[VAR_ADDR:%.+]] = alloca [[S_FLOAT_TY]]*, |
227 | // CHECK: [[SVAR_ADDR:%.+]] = alloca i{{[0-9]+}}*, |
228 | // CHECK: [[TMP:%.+]] = alloca [[S_FLOAT_TY]]*, |
229 | |
230 | // discard omp loop variables |
231 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
232 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
233 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
234 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
235 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
236 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
237 | |
238 | // CHECK-DAG: [[T_VAR_PRIV:%.+]] = alloca i{{[0-9]+}}, |
239 | // CHECK-DAG: [[VEC_PRIV:%.+]] = alloca [2 x i{{[0-9]+}}], |
240 | // CHECK-DAG: [[S_ARR_PRIV:%.+]] = alloca [2 x [[S_FLOAT_TY]]], |
241 | // CHECK-DAG: [[VAR_PRIV:%.+]] = alloca [[S_FLOAT_TY]], |
242 | // CHECK-DAG: [[TMP_PRIV:%.+]] = alloca [[S_FLOAT_TY]]*, |
243 | // CHECK: [[SVAR_PRIV:%.+]] = alloca i{{[0-9]+}}, |
244 | |
245 | // CHECK: store i{{[0-9]+}}* [[T_VAR_IN]], i{{[0-9]+}}** [[T_VAR_ADDR]], |
246 | // CHECK: store [2 x i{{[0-9]+}}]* [[VEC_IN]], [2 x i{{[0-9]+}}]** [[VEC_ADDR]], |
247 | // CHECK: store [2 x [[S_FLOAT_TY]]]* [[S_ARR_IN]], [2 x [[S_FLOAT_TY]]]** [[S_ARR_ADDR]], |
248 | // CHECK: store [[S_FLOAT_TY]]* [[VAR_IN]], [[S_FLOAT_TY]]** [[VAR_ADDR]], |
249 | // CHECK: store i{{[0-9]+}}* [[SVAR_IN]], i{{[0-9]+}}** [[SVAR_ADDR]], |
250 | |
251 | // init t_var |
252 | // CHECK-DAG: [[T_VAR_ADDR_CONV_VAL_REF:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[T_VAR_ADDR]], |
253 | // CHECK-DAG: [[T_VAR_ADDR_CONV_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[T_VAR_ADDR_CONV_VAL_REF]], |
254 | // CHECK-DAG: store i{{[0-9]+}} [[T_VAR_ADDR_CONV_VAL]], i{{[0-9]+}}* [[T_VAR_PRIV]], |
255 | |
256 | // init vec |
257 | // CHECK-DAG: [[VEC_ADDR_VAL:%.+]] = load [2 x i{{[0-9]+}}]*, [2 x i{{[0-9]+}}]** [[VEC_ADDR]], |
258 | // CHECK-DAG: [[VEC_ADDR_VAL_BCAST:%.+]] = bitcast [2 x i{{[0-9]+}}]* [[VEC_ADDR_VAL]] to i{{[0-9]+}}* |
259 | // CHECK-DAG: [[VEC_PRIV_BCAST:%.+]] = bitcast [2 x i{{[0-9]+}}]* [[VEC_PRIV]] to i{{[0-9]+}}* |
260 | // CHECK-DAG: call void @llvm.memcpy.{{.*}}(i{{[0-9]+}}* align {{[0-9]+}} [[VEC_PRIV_BCAST]], i{{[0-9]+}}* align {{[0-9]+}} [[VEC_ADDR_VAL_BCAST]],{{.+}}) |
261 | |
262 | // init s_arr |
263 | // CHECK-DAG: [[S_ARR_ADDR_VAL:%.+]] = load [2 x [[S_FLOAT_TY]]]*, [2 x [[S_FLOAT_TY]]]** [[S_ARR_ADDR]], |
264 | // CHECK-DAG: [[S_ARR_ADDR_BCAST:%.+]] = bitcast [2 x [[S_FLOAT_TY]]]* [[S_ARR_ADDR_VAL]] to [[S_FLOAT_TY]]* |
265 | // CHECK-DAG: [[S_ARR_PRIV_BGN:%.+]] = getelementptr{{.+}} [2 x [[S_FLOAT_TY]]], [2 x [[S_FLOAT_TY]]]* [[S_ARR_PRIV]]{{.+}} |
266 | // CHECK-DAG: [[S_ARR_PRIV_NEXT:%.+]] = getelementptr [[S_FLOAT_TY]], [[S_FLOAT_TY]]* [[S_ARR_PRIV_BGN]]{{.+}} |
267 | // CHECK-DAG: [[S_ARR_IS_EMPTY:%.+]] = icmp eq [[S_FLOAT_TY]]* [[S_ARR_PRIV_BGN]], [[S_ARR_PRIV_NEXT]] |
268 | // CHECK-DAG: br i1 [[S_ARR_IS_EMPTY]], label %[[S_ARR_CPY_DONE:.+]], label %[[S_ARR_CPY_BODY:.+]] |
269 | |
270 | // CHECK-DAG: [[S_ARR_CPY_BODY]]: |
271 | // CHECK-DAG: [[S_ARR_SRC_PAST:%.+]] = phi{{.+}} [ [[S_ARR_ADDR_BCAST]],{{.+}} ], [ [[S_ARR_SRC:%.+]],{{.+}} ] |
272 | // CHECK-DAG: [[S_ARR_DST_PAST:%.+]] = phi{{.+}} [ [[S_ARR_PRIV_BGN]],{{.+}} ], [ [[S_ARR_DST:%.+]],{{.+}} ] |
273 | // CHECK-DAG: [[S_ARR_SRC_BCAST:%.+]] = bitcast{{.+}} [[S_ARR_SRC_PAST]] to{{.+}} |
274 | // CHECK-DAG: [[S_ARR_DST_BCAST:%.+]] = bitcast{{.+}} [[S_ARR_DST_PAST]] to{{.+}} |
275 | // CHECK-DAG: call{{.+}} @llvm.memcpy.{{.+}}({{.+}}* align {{[0-9]+}} [[S_ARR_DST_BCAST]], {{.+}}* align {{[0-9]+}} [[S_ARR_SRC_BCAST]]{{.+}}) |
276 | // CHECK-DAG: [[S_ARR_SRC]] = getelementptr{{.+}} |
277 | // CHECK-DAG: [[S_ARR_DST]] = getelementptr{{.+}} |
278 | // CHECK-DAG: [[S_ARR_CPY_FIN:%.+]] = icmp{{.+}} [[S_ARR_DST]], [[S_ARR_PRIV_NEXT]] |
279 | // CHECK-DAG: br i1 [[S_ARR_CPY_FIN]], label %[[S_ARR_CPY_DONE]], label %[[S_ARR_CPY_BODY]] |
280 | // CHECK-DAG: [[S_ARR_CPY_DONE]]: |
281 | |
282 | // init var |
283 | // CHECK-DAG: [[VAR_ADDR_VAL:%.+]] = load [[S_FLOAT_TY]]*, [[S_FLOAT_TY]]** [[VAR_ADDR]], |
284 | // CHECK-DAG: store{{.+}} [[VAR_ADDR_VAL]],{{.+}} [[TMP]], |
285 | // CHECK-DAG: [[TMP_VAL:%.+]] = load [[S_FLOAT_TY]]*, [[S_FLOAT_TY]]** [[TMP]], |
286 | // CHECK-DAG: [[VAR_PRIV_BCAST:%.+]] = bitcast [[S_FLOAT_TY]]* [[VAR_PRIV]] to{{.+}} |
287 | // CHECK-DAG: [[TMP_BCAST:%.+]] = bitcast [[S_FLOAT_TY]]* [[TMP_VAL]] to{{.+}} |
288 | // CHECK-DAG: call{{.+}} @llvm.memcpy.{{.+}}({{.+}}* align {{[0-9]+}} [[VAR_PRIV_BCAST]], {{.+}}* align {{[0-9]+}} [[TMP_BCAST]],{{.+}}) |
289 | // CHECK-DAG: store [[S_FLOAT_TY]]* [[VAR_PRIV]], [[S_FLOAT_TY]]** [[TMP_PRIV]], |
290 | |
291 | // init svar |
292 | // CHECK-DAG: [[SVAR_CONV_VAL_REF:%.+]] = load{{.+}},{{.+}} [[SVAR_ADDR]], |
293 | // CHECK-DAG: [[SVAR_CONV_VAL:%.+]] = load{{.+}},{{.+}} [[SVAR_CONV_VAL_REF]], |
294 | // CHECK-DAG: store{{.+}} [[SVAR_CONV_VAL]],{{.+}} [[SVAR_PRIV]], |
295 | |
296 | // CHECK-DAG: store i{{[0-9]+}} 0, i{{[0-9]+}}* %.omp{{.+}}, |
297 | // CHECK-DAG: store i{{[0-9]+}} 1, i{{[0-9]+}}* %.omp{{.+}}, |
298 | // CHECK-DAG: store i{{[0-9]+}} 1, i{{[0-9]+}}* %.omp{{.+}}, |
299 | // CHECK-DAG: store i{{[0-9]+}} 0, i{{[0-9]+}}* %.omp{{.+}}, |
300 | |
301 | // CHECK: call void @__kmpc_for_static_init_4( |
302 | // CHECK: call void @__kmpc_for_static_fini( |
303 | // CHECK: ret void |
304 | |
305 | // Template |
306 | // CHECK: define{{.*}} i{{[0-9]+}} [[TMAIN_INT:@.+]]() |
307 | // CHECK: [[TEST:%.+]] = alloca [[S_INT_TY]], |
308 | // CHECK: call {{.*}} [[S_INT_TY_DEF_CONSTR:@.+]]([[S_INT_TY]]* [[TEST]]) |
309 | // CHECK: call i{{[0-9]+}} @__tgt_target_teams( |
310 | // CHECK: call void [[OFFLOAD_FUN_1:@.+]]( |
311 | // CHECK: ret |
312 | |
313 | // CHECK: define{{.+}} [[OFFLOAD_FUN_1]]( |
314 | // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 4, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*, i{{[0-9]+}}*, [2 x i{{[0-9]+}}]*, [2 x [[S_INT_TY]]]*, [[S_INT_TY]]*)* [[OMP_OUTLINED_1:@.+]] to void |
315 | // CHECK: ret |
316 | // |
317 | // CHECK: define internal void [[OMP_OUTLINED_1]](i{{[0-9]+}}*{{.+}}, i{{[0-9]+}}*{{.+}}, i{{[0-9]+}}*{{.+}} [[T_VAR_IN:%.+]], [2 x i{{[0-9]+}}]*{{.*}} [[VEC_IN:%.+]], [2 x [[S_INT_TY]]]*{{.*}} [[S_ARR_IN:%.+]], [[S_INT_TY]]*{{.*}} [[VAR_IN:%.+]]) |
318 | |
319 | // CHECK: alloca i{{[0-9]+}}*, |
320 | // CHECK: alloca i{{[0-9]+}}*, |
321 | // CHECK: [[T_VAR_ADDR:%.+]] = alloca i{{[0-9]+}}*, |
322 | // CHECK: [[VEC_ADDR:%.+]] = alloca [2 x i{{[0-9]+}}]*, |
323 | // CHECK: [[S_ARR_ADDR:%.+]] = alloca [2 x [[S_INT_TY]]]*, |
324 | // CHECK: [[VAR_ADDR:%.+]] = alloca [[S_INT_TY]]*, |
325 | // CHECK: [[TMP:%.+]] = alloca [[S_INT_TY]]*, |
326 | |
327 | // discard omp loop variables |
328 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
329 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
330 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
331 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
332 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
333 | // CHECK: {{.*}} = alloca i{{[0-9]+}}, |
334 | |
335 | // CHECK-DAG: [[T_VAR_PRIV:%.+]] = alloca i{{[0-9]+}}, |
336 | // CHECK-DAG: [[VEC_PRIV:%.+]] = alloca [2 x i{{[0-9]+}}], |
337 | // CHECK-DAG: [[S_ARR_PRIV:%.+]] = alloca [2 x [[S_INT_TY]]], |
338 | // CHECK-DAG: [[VAR_PRIV:%.+]] = alloca [[S_INT_TY]], |
339 | // CHECK-DAG: [[TMP_PRIV:%.+]] = alloca [[S_INT_TY]]*, |
340 | |
341 | // CHECK: store i{{[0-9]+}}* [[T_VAR_IN]], i{{[0-9]+}}** [[T_VAR_ADDR]], |
342 | // CHECK: store [2 x i{{[0-9]+}}]* [[VEC_IN]], [2 x i{{[0-9]+}}]** [[VEC_ADDR]], |
343 | // CHECK: store [2 x [[S_INT_TY]]]* [[S_ARR_IN]], [2 x [[S_INT_TY]]]** [[S_ARR_ADDR]], |
344 | // CHECK: store [[S_INT_TY]]* [[VAR_IN]], [[S_INT_TY]]** [[VAR_ADDR]], |
345 | |
346 | // init t_var |
347 | // CHECK-DAG: [[T_VAR_ADDR_CONV_VAL_REF:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[T_VAR_ADDR]], |
348 | // CHECK-DAG: [[T_VAR_ADDR_CONV_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[T_VAR_ADDR_CONV_VAL_REF]], |
349 | // CHECK-DAG: store i{{[0-9]+}} [[T_VAR_ADDR_CONV_VAL]], i{{[0-9]+}}* [[T_VAR_PRIV]], |
350 | |
351 | // init vec |
352 | // CHECK-DAG: [[VEC_ADDR_VAL:%.+]] = load [2 x i{{[0-9]+}}]*, [2 x i{{[0-9]+}}]** [[VEC_ADDR]], |
353 | // CHECK-DAG: [[VEC_ADDR_VAL_BCAST:%.+]] = bitcast [2 x i{{[0-9]+}}]* [[VEC_ADDR_VAL]] to i{{[0-9]+}}* |
354 | // CHECK-DAG: [[VEC_PRIV_BCAST:%.+]] = bitcast [2 x i{{[0-9]+}}]* [[VEC_PRIV]] to i{{[0-9]+}}* |
355 | // CHECK-DAG: call void @llvm.memcpy.{{.*}}(i{{[0-9]+}}* align {{[0-9]+}} [[VEC_PRIV_BCAST]], i{{[0-9]+}}* align {{[0-9]+}} [[VEC_ADDR_VAL_BCAST]],{{.+}}) |
356 | |
357 | // init s_arr |
358 | // CHECK-DAG: [[S_ARR_ADDR_VAL:%.+]] = load [2 x [[S_INT_TY]]]*, [2 x [[S_INT_TY]]]** [[S_ARR_ADDR]], |
359 | // CHECK-DAG: [[S_ARR_ADDR_BCAST:%.+]] = bitcast [2 x [[S_INT_TY]]]* [[S_ARR_ADDR_VAL]] to [[S_INT_TY]]* |
360 | // CHECK-DAG: [[S_ARR_PRIV_BGN:%.+]] = getelementptr{{.+}} [2 x [[S_INT_TY]]], [2 x [[S_INT_TY]]]* [[S_ARR_PRIV]]{{.+}} |
361 | // CHECK-DAG: [[S_ARR_PRIV_NEXT:%.+]] = getelementptr [[S_INT_TY]], [[S_INT_TY]]* [[S_ARR_PRIV_BGN]]{{.+}} |
362 | // CHECK-DAG: [[S_ARR_IS_EMPTY:%.+]] = icmp eq [[S_INT_TY]]* [[S_ARR_PRIV_BGN]], [[S_ARR_PRIV_NEXT]] |
363 | // CHECK-DAG: br i1 [[S_ARR_IS_EMPTY]], label %[[S_ARR_CPY_DONE:.+]], label %[[S_ARR_CPY_BODY:.+]] |
364 | |
365 | // CHECK-DAG: [[S_ARR_CPY_BODY]]: |
366 | // CHECK-DAG: [[S_ARR_SRC_PAST:%.+]] = phi{{.+}} [ [[S_ARR_ADDR_BCAST]],{{.+}} ], [ [[S_ARR_SRC:%.+]],{{.+}} ] |
367 | // CHECK-DAG: [[S_ARR_DST_PAST:%.+]] = phi{{.+}} [ [[S_ARR_PRIV_BGN]],{{.+}} ], [ [[S_ARR_DST:%.+]],{{.+}} ] |
368 | // CHECK-DAG: [[S_ARR_SRC_BCAST:%.+]] = bitcast{{.+}} [[S_ARR_SRC_PAST]] to{{.+}} |
369 | // CHECK-DAG: [[S_ARR_DST_BCAST:%.+]] = bitcast{{.+}} [[S_ARR_DST_PAST]] to{{.+}} |
370 | // CHECK-DAG: call{{.+}} @llvm.memcpy.{{.+}}({{.+}}* align {{[0-9]+}} [[S_ARR_DST_BCAST]], {{.+}}* align {{[0-9]+}} [[S_ARR_SRC_BCAST]]{{.+}}) |
371 | // CHECK-DAG: [[S_ARR_SRC]] = getelementptr{{.+}} |
372 | // CHECK-DAG: [[S_ARR_DST]] = getelementptr{{.+}} |
373 | // CHECK-DAG: [[S_ARR_CPY_FIN:%.+]] = icmp{{.+}} [[S_ARR_DST]], [[S_ARR_PRIV_NEXT]] |
374 | // CHECK-DAG: br i1 [[S_ARR_CPY_FIN]], label %[[S_ARR_CPY_DONE]], label %[[S_ARR_CPY_BODY]] |
375 | // CHECK-DAG: [[S_ARR_CPY_DONE]]: |
376 | |
377 | // init var |
378 | // CHECK-DAG: [[VAR_ADDR_VAL:%.+]] = load [[S_INT_TY]]*, [[S_INT_TY]]** [[VAR_ADDR]], |
379 | // CHECK-DAG: store{{.+}} [[VAR_ADDR_VAL]],{{.+}} [[TMP]], |
380 | // CHECK-DAG: [[TMP_VAL:%.+]] = load [[S_INT_TY]]*, [[S_INT_TY]]** [[TMP]], |
381 | // CHECK-DAG: [[VAR_PRIV_BCAST:%.+]] = bitcast [[S_INT_TY]]* [[VAR_PRIV]] to{{.+}} |
382 | // CHECK-DAG: [[TMP_BCAST:%.+]] = bitcast [[S_INT_TY]]* [[TMP_VAL]] to{{.+}} |
383 | // CHECK-DAG: call{{.+}} @llvm.memcpy.{{.+}}({{.+}}* align {{[0-9]+}} [[VAR_PRIV_BCAST]], {{.+}}* align {{[0-9]+}} [[TMP_BCAST]],{{.+}}) |
384 | // CHECK-DAG: store [[S_INT_TY]]* [[VAR_PRIV]], [[S_INT_TY]]** [[TMP_PRIV]], |
385 | |
386 | // CHECK-DAG: store i{{[0-9]+}} 0, i{{[0-9]+}}* %.omp{{.+}}, |
387 | // CHECK-DAG: store i{{[0-9]+}} 1, i{{[0-9]+}}* %.omp{{.+}}, |
388 | // CHECK-DAG: store i{{[0-9]+}} 1, i{{[0-9]+}}* %.omp{{.+}}, |
389 | // CHECK-DAG: store i{{[0-9]+}} 0, i{{[0-9]+}}* %.omp{{.+}}, |
390 | |
391 | // CHECK: call void @__kmpc_for_static_init_4( |
392 | // CHECK: call void @__kmpc_for_static_fini( |
393 | // CHECK: ret void |
394 | |
395 | // CHECK: !{!"llvm.loop.vectorize.enable", i1 true} |
396 | #endif |
397 | |