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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | 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 - | FileCheck --check-prefix SIMD-ONLY1 %s |
29 | // SIMD-ONLY1-NOT: {{__kmpc|__tgt}} |
30 | // expected-no-diagnostics |
31 | #ifndef HEADER |
32 | #define HEADER |
33 | template <class T> |
34 | struct S { |
35 | T f; |
36 | S(T a) : f(a) {} |
37 | S() : f() {} |
38 | operator T() { return T(); } |
39 | ~S() {} |
40 | }; |
41 | |
42 | volatile int g __attribute__((aligned(128))) = 1212; |
43 | |
44 | struct SS { |
45 | int a; |
46 | int b : 4; |
47 | int &c; |
48 | SS(int &d) : a(0), b(0), c(d) { |
49 | #pragma omp target |
50 | #pragma omp teams private(a, b, c) |
51 | #ifdef LAMBDA |
52 | [&]() { |
53 | ++this->a, --b, (this)->c /= 1; |
54 | }(); |
55 | #else |
56 | ++this->a, --b, c /= 1; |
57 | #endif |
58 | } |
59 | }; |
60 | |
61 | template<typename T> |
62 | struct SST { |
63 | T a; |
64 | SST() : a(T()) { |
65 | #pragma omp target |
66 | #pragma omp teams private(a) |
67 | #ifdef LAMBDA |
68 | [&]() { |
69 | [&]() { |
70 | ++this->a; |
71 | }(); |
72 | }(); |
73 | #else |
74 | ++(this)->a; |
75 | #endif |
76 | } |
77 | }; |
78 | |
79 | // CHECK: [[SS_TY:%.+]] = type { i{{[0-9]+}}, i8 |
80 | // LAMBDA: [[SS_TY:%.+]] = type { i{{[0-9]+}}, i8 |
81 | // LAMBDA: [[CAP_0_TY:%.+]] = type { [[SS_TY]]*, i{{[0-9]+}}*, |
82 | // LAMBDA: [[CAP_1_TY:%.+]] = type { i{{[0-9]+}}*, i{{[0-9]+}}* } |
83 | // CHECK: [[S_FLOAT_TY:%.+]] = type { float } |
84 | // CHECK: [[S_INT_TY:%.+]] = type { i{{[0-9]+}} } |
85 | // CHECK: [[SST_TY:%.+]] = type { i{{[0-9]+}} } |
86 | template <typename T> |
87 | T tmain() { |
88 | S<T> test; |
89 | SST<T> sst; |
90 | T t_var __attribute__((aligned(128))) = T(); |
91 | T vec[] __attribute__((aligned(128))) = {1, 2}; |
92 | S<T> s_arr[] __attribute__((aligned(128))) = {1, 2}; |
93 | S<T> var __attribute__((aligned(128))) (3); |
94 | #pragma omp target |
95 | #pragma omp teams private(t_var, vec, s_arr, var) |
96 | { |
97 | vec[0] = t_var; |
98 | s_arr[0] = var; |
99 | } |
100 | return T(); |
101 | } |
102 | |
103 | int main() { |
104 | static int sivar; |
105 | SS ss(sivar); |
106 | #ifdef LAMBDA |
107 | // LAMBDA: [[G:@.+]] = global i{{[0-9]+}} 1212, |
108 | // LAMBDA: define {{.+}} @main() |
109 | // LAMBDA: alloca [[SS_TY]], |
110 | // LAMBDA: alloca [[CAP_TY:%.+]], |
111 | |
112 | // LAMBDA: call{{.*}} [[ST_CONSTR_INIT:@.+]]([[SS_TY]]* |
113 | // LAMBDA: call{{.*}} void [[OUTER_LAMBDA:@[^(]+]]([[CAP_TY]]* |
114 | |
115 | // lambda and target region in main |
116 | // LAMBDA: define {{.+}} [[OUTER_LAMBDA]]([[CAP_TY]]* {{.+}}) |
117 | // LAMBDA: call void @[[OMP_OFFLOADING:.+]]() |
118 | |
119 | // target region in struct constructor |
120 | // LAMBDA: define{{.*}} void [[ST_CONSTR:@.+]]([[SS_TY]]* %this, |
121 | // LAMBDA: call void [[OMP_OFFLOADING_1:@.+]]([[SS_TY]] |
122 | |
123 | // offloading function in struct constructor |
124 | // LAMBDA: define{{.*}} void [[OMP_OFFLOADING_1]]([[SS_TY]] |
125 | // LAMBDA: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 1, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*, [[SS_TY]]*)* [[OMP_OUTLINED:@.+]] to void |
126 | |
127 | // outlined teams region in struct constructor |
128 | // LAMBDA: define{{.*}} void [[OMP_OUTLINED]](i{{[0-9]+}}* {{.+}}, i{{[0-9]+}}* {{.+}}, [[SS_TY]]* |
129 | // LAMBDA: [[THIS_ADDR:%.+]] = alloca [[SS_TY]]*, |
130 | // LAMBDA: [[A_PRIV:%.+]] = alloca i{{[0-9]+}}, |
131 | // LAMBDA: [[B_PRIV:%.+]] = alloca i{{[0-9]+}}, |
132 | // LAMBDA: [[C_PRIV:%.+]] = alloca i{{[0-9]+}}, |
133 | // LAMBDA: [[THIS_REF:%.+]] = load [[SS_TY]]*, [[SS_TY]]** [[THIS_ADDR]], |
134 | // LAMBDA: store i{{[0-9]+}}* [[A_PRIV]], i{{[0-9]+}}** [[A_TMP_REF:%.+]], |
135 | // LAMBDA: store i{{[0-9]+}}* [[C_PRIV]], i{{[0-9]+}}** [[C_TMP_REF:%.+]], |
136 | // LAMBDA: [[CAP_THIS_REF:%.+]] = getelementptr {{.+}} [[CAP_0_TY]], [[CAP_0_TY]]* {{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 0 |
137 | // LAMBDA: store [[SS_TY]]* [[THIS_REF]], [[SS_TY]]** [[CAP_THIS_REF]], |
138 | // LAMBDA: [[CAP_A_REF:%.+]] = getelementptr {{.+}} [[CAP_0_TY]], [[CAP_0_TY]]* {{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 1 |
139 | // LAMBDA: [[A_TMP_VAL:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[A_TMP_REF]], |
140 | // LAMBDA: store i{{[0-9]+}}* [[A_TMP_VAL]], i{{[0-9]+}}** [[CAP_A_REF]], |
141 | // LAMBDA: [[CAP_B_REF:%.+]] = getelementptr {{.+}} [[CAP_0_TY]], [[CAP_0_TY]]* {{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 2 |
142 | // LAMBDA: store i{{[0-9]+}}* [[B_PRIV]], i{{[0-9]+}}** [[CAP_B_REF]], |
143 | // LAMBDA: [[CAP_C_REF:%.+]] = getelementptr {{.+}} [[CAP_0_TY]], [[CAP_0_TY]]* {{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 3 |
144 | // LAMBDA: [[C_TMP_VAL:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[C_TMP_REF]], |
145 | // LAMBDA: store i{{[0-9]+}}* [[C_TMP_VAL]], i{{[0-9]+}}** [[CAP_C_REF]], |
146 | // call void [[INNER_LAMBDA_CONSTR:@.+]]([[CAP_0_TY]]* |
147 | |
148 | // inner lambda in struct constructor |
149 | // define{{.*}} void [[INNER_LAMBDA_CONSTR]]([[CAP_0_TY]]* |
150 | // LAMBDA: [[CAP_A_REF_1:%.+]] = getelementptr {{.+}} [[CAP_0_TY]], [[CAP_0_TY]]* {{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 1 |
151 | // LAMBDA: [[A_REF_FROM_CAP:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[CAP_A_REF_1]], |
152 | // LAMBDA: [[A_VAL_FROM_CAP:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[A_REF_FROM_CAP]], |
153 | // LAMBDA: [[A_INC_VAL:%.+]] = add {{.+}} i{{[0-9]+}} [[A_VAL_FROM_CAP]], 1 |
154 | // LAMBDA: store i{{[0-9]+}} [[A_INC_VAL]], i{{[0-9]+}}* [[A_REF_FROM_CAP]], |
155 | |
156 | // LAMBDA: [[CAP_B_REF_1:%.+]] = getelementptr {{.+}} [[CAP_0_TY]], [[CAP_0_TY]]* {{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 2 |
157 | // LAMBDA: [[B_REF_FROM_CAP:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[CAP_B_REF_1]], |
158 | // LAMBDA: [[B_VAL_FROM_CAP:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[B_REF_FROM_CAP]], |
159 | // LAMBDA: [[B_DEC_VAL:%.+]] = add {{.+}} i{{[0-9]+}} [[B_VAL_FROM_CAP]], -1 |
160 | // LAMBDA: store i{{[0-9]+}} [[B_DEC_VAL]], i{{[0-9]+}}* [[B_REF_FROM_CAP]], |
161 | |
162 | // LAMBDA: [[CAP_C_REF_1:%.+]] = getelementptr {{.+}} [[CAP_0_TY]], [[CAP_0_TY]]* {{.+}}, i{{[0-9]+}} 0, i{{[0-9]+}} 3 |
163 | // LAMBDA: [[C_REF_FROM_CAP:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[CAP_C_REF_1]], |
164 | // LAMBDA: [[C_VAL_FROM_CAP:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[C_REF_FROM_CAP]], |
165 | // LAMBDA: [[C_DEC_VAL:%.+]] = sdiv{{.*}} i{{[0-9]+}} [[C_VAL_FROM_CAP]], 1 |
166 | // LAMBDA: store i{{[0-9]+}} [[C_DEC_VAL]], i{{[0-9]+}}* [[C_REF_FROM_CAP]], |
167 | // ret |
168 | |
169 | [&]() { |
170 | #pragma omp target |
171 | #pragma omp teams private(g, sivar) |
172 | { |
173 | // LAMBDA: define{{.+}} @[[OMP_OFFLOADING]]() |
174 | // LAMBDA: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 0, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*)* [[OMP_OUTLINED_1:@.+]] to void |
175 | |
176 | // LAMBDA: define {{.+}} [[OMP_OUTLINED_1]](i{{[0-9]+}}* {{.+}}, i{{[0-9]+}}* {{.+}} |
177 | // LAMBDA: [[G_LOC_OUTER:%.+]] = alloca i{{[0-9]+}}, |
178 | // LAMBDA: [[SIVAR_LOC_OUTER:%.+]] = alloca i{{[0-9]+}}, |
179 | // LAMBDA: store i{{[0-9]+}} 1, i{{[0-9]+}}* [[G_LOC_OUTER]] |
180 | // LAMBDA: store i{{[0-9]+}} 2, i{{[0-9]+}}* [[SIVAR_LOC_OUTER]] |
181 | // LAMBDA: call{{.*}} void [[INNER_LAMBDA:@[^(]+]]([[CAP_1_TY]]* |
182 | // LAMBDA: ret |
183 | g = 1; |
184 | sivar = 2; |
185 | [&]() { |
186 | // LAMBDA: define {{.+}} [[INNER_LAMBDA]]([[CAP_1_TY]]* {{.+}}) |
187 | g = 2; |
188 | sivar = 4; |
189 | // LAMBDA: store i{{[0-9]+}} 2, i{{[0-9]+}}* |
190 | // LAMBDA: store i{{[0-9]+}} 4, i{{[0-9]+}}* |
191 | }(); |
192 | } |
193 | }(); |
194 | return 0; |
195 | #else |
196 | S<float> test; |
197 | int t_var = 0; |
198 | int vec[] = {1, 2}; |
199 | S<float> s_arr[] = {1, 2}; |
200 | S<float> var(3); |
201 | #pragma omp target |
202 | #pragma omp teams private(t_var, vec, s_arr, var, sivar) |
203 | { |
204 | vec[0] = t_var; |
205 | s_arr[0] = var; |
206 | sivar = 3; |
207 | } |
208 | return tmain<int>(); |
209 | #endif |
210 | } |
211 | |
212 | // CHECK: define{{.*}} i{{[0-9]+}} @main() |
213 | // CHECK: [[TEST:%.+]] = alloca [[S_FLOAT_TY]], |
214 | // CHECK: call {{.*}} [[S_FLOAT_TY_DEF_CONSTR:@.+]]([[S_FLOAT_TY]]* [[TEST]]) |
215 | // CHECK: call void @[[OMP_OFFLOADING:.+]]() |
216 | // CHECK: = call{{.*}} i{{.+}} [[TMAIN_INT:@.+]]() |
217 | // CHECK: call void [[S_FLOAT_TY_DESTR:@.+]]([[S_FLOAT_TY]]* |
218 | // CHECK: ret |
219 | |
220 | // target region in main function |
221 | // CHECK: define{{.+}} @[[OMP_OFFLOADING]]() |
222 | // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 0, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*)* [[OMP_OUTLINED:@.+]] to void |
223 | |
224 | // CHECK: define internal void [[OMP_OUTLINED]](i{{[0-9]+}}* noalias [[GTID_ADDR:%.+]], i{{[0-9]+}}* noalias %{{.+}}) |
225 | // CHECK: [[T_VAR_PRIV:%.+]] = alloca i{{[0-9]+}}, |
226 | // CHECK: [[VEC_PRIV:%.+]] = alloca [2 x i{{[0-9]+}}], |
227 | // CHECK: [[S_ARR_PRIV:%.+]] = alloca [2 x [[S_FLOAT_TY]]], |
228 | // CHECK: [[VAR_PRIV:%.+]] = alloca [[S_FLOAT_TY]], |
229 | // CHECK: [[SIVAR_PRIV:%.+]] = alloca i{{[0-9]+}}, |
230 | // CHECK: store i{{[0-9]+}}* [[GTID_ADDR]], i{{[0-9]+}}** [[GTID_ADDR_REF:%.+]] |
231 | // CHECK-NOT: [[T_VAR_PRIV]] |
232 | // CHECK-NOT: [[VEC_PRIV]] |
233 | // CHECK: {{.+}}: |
234 | // CHECK: [[S_ARR_PRIV_ITEM:%.+]] = phi [[S_FLOAT_TY]]* |
235 | // CHECK: call {{.*}} [[S_FLOAT_TY_DEF_CONSTR]]([[S_FLOAT_TY]]* [[S_ARR_PRIV_ITEM]]) |
236 | // CHECK-NOT: [[T_VAR_PRIV]] |
237 | // CHECK-NOT: [[VEC_PRIV]] |
238 | // CHECK: call {{.*}} [[S_FLOAT_TY_DEF_CONSTR]]([[S_FLOAT_TY]]* [[VAR_PRIV]]) |
239 | // CHECK-DAG: call void [[S_FLOAT_TY_DESTR]]([[S_FLOAT_TY]]* [[VAR_PRIV]]) |
240 | // CHECK-DAG: call void [[S_FLOAT_TY_DESTR]]([[S_FLOAT_TY]]* |
241 | // CHECK: ret void |
242 | |
243 | // template tmain |
244 | // CHECK: define{{.*}} i{{[0-9]+}} [[TMAIN_INT]]() |
245 | // CHECK: [[TEST:%.+]] = alloca [[S_INT_TY]], |
246 | // CHECK: call {{.*}} [[S_INT_TY_DEF_CONSTR:@.+]]([[S_INT_TY]]* [[TEST]]) |
247 | // CHECK: call void [[S_INT_TY_CONSTR:@.+]]([[S_INT_TY]]* {{.+}}, i{{[0-9]+}}{{.*}} 3) |
248 | // CHECK: call void [[OMP_OFFLOADING_TMAIN:@.+]]() |
249 | |
250 | // target in SS constructor |
251 | // CHECK: define{{.+}} [[OMP_OFFLOADING_SS:@.+]]([[SS_TY]]* |
252 | // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 1, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*, [[SS_TY]]*)* [[OMP_OUTLINED_SS:@.+]] to void |
253 | |
254 | // CHECK: define{{.*}} void [[OMP_OUTLINED_SS]](i{{[0-9]+}}* {{.+}}, i{{[0-9]+}}* {{.+}}, [[SS_TY]]* |
255 | // CHECK: [[A_PRIV:%.+]] = alloca i{{[0-9]+}}, |
256 | // CHECK: [[B_PRIV:%.+]] = alloca i{{[0-9]+}}, |
257 | // CHECK: [[C_PRIV:%.+]] = alloca i{{[0-9]+}}, |
258 | // CHECK: store i{{[0-9]+}}* [[A_PRIV]], i{{[0-9]+}}** [[A_REF:%.+]], |
259 | // CHECK: store i{{[0-9]+}}* [[C_PRIV]], i{{[0-9]+}}** [[C_REF:%.+]], |
260 | // CHECK: [[A_REF_VAL:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[A_REF]] |
261 | // CHECK: [[A_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[A_REF_VAL]] |
262 | // CHECK: [[A_INC:%.+]] = add{{.*}} i{{[0-9]+}} [[A_VAL]], 1 |
263 | // CHECK: store i{{[0-9]+}} [[A_INC]], i{{[0-9]+}}* [[A_REF_VAL]], |
264 | // CHECK: [[B_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[B_PRIV]] |
265 | // CHECK: [[B_DEC:%.+]] = add{{.*}} i{{[0-9]+}} [[B_VAL]], -1 |
266 | // CHECK: store i{{[0-9]+}} [[B_DEC]], i{{[0-9]+}}* [[B_PRIV]], |
267 | // CHECK: [[C_REF_VAL:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[C_REF]] |
268 | // CHECK: [[C_VAL:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[C_REF_VAL]] |
269 | // CHECK: [[C_DIV:%.+]] = sdiv i{{[0-9]+}} [[C_VAL]], 1 |
270 | // CHECK: store i{{[0-9]+}} [[C_DIV]], i{{[0-9]+}}* [[C_REF_VAL]], |
271 | // CHECK: ret |
272 | |
273 | // target in tmain template |
274 | // CHECK: define{{.+}} [[OMP_OFFLOADING_TMAIN]]() |
275 | // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 0, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*)* [[OMP_OUTLINED_TMAIN:@.+]] to void |
276 | |
277 | // CHECK: define{{.*}} void [[OMP_OUTLINED_TMAIN]](i{{[0-9]+}}* noalias [[GTID_ADDR:%.+]], i{{[0-9]+}}* noalias %{{.+}}) |
278 | // CHECK: [[T_VAR_PRIV:%.+]] = alloca i{{[0-9]+}}, align 128 |
279 | // CHECK: [[VEC_PRIV:%.+]] = alloca [2 x i{{[0-9]+}}], align 128 |
280 | // CHECK: [[S_ARR_PRIV:%.+]] = alloca [2 x [[S_INT_TY]]], align 128 |
281 | // CHECK: [[VAR_PRIV:%.+]] = alloca [[S_INT_TY]], align 128 |
282 | // CHECK: store i{{[0-9]+}}* [[GTID_ADDR]], i{{[0-9]+}}** [[GTID_ADDR_REF:%.+]] |
283 | // CHECK-NOT: [[T_VAR_PRIV]] |
284 | // CHECK-NOT: [[VEC_PRIV]] |
285 | // CHECK: {{.+}}: |
286 | // CHECK: [[S_ARR_PRIV_ITEM:%.+]] = phi [[S_INT_TY]]* |
287 | // CHECK: call {{.*}} [[S_INT_TY_DEF_CONSTR]]([[S_INT_TY]]* [[S_ARR_PRIV_ITEM]]) |
288 | // CHECK-NOT: [[T_VAR_PRIV]] |
289 | // CHECK-NOT: [[VEC_PRIV]] |
290 | // CHECK: call {{.*}} [[S_INT_TY_DEF_CONSTR]]([[S_INT_TY]]* [[VAR_PRIV]]) |
291 | // CHECK-DAG: call void [[S_INT_TY_DESTR:@.+]]([[S_INT_TY]]* [[VAR_PRIV]]) |
292 | // CHECK-DAG: call void [[S_INT_TY_DESTR]]([[S_INT_TY]]* |
293 | // CHECK: ret |
294 | |
295 | // SST constructor |
296 | // CHECK: define{{.+}} [[SST_CONST:@.+]]([[SST_TY]]* {{.+}}) |
297 | // CHECK: call void [[OMP_OFFLOADING_SST:@.+]]([[SST_TY]]* {{.+}}) |
298 | |
299 | // target in SST constructor |
300 | // CHECK: define{{.+}} [[OMP_OFFLOADING_SST]]([[SST_TY]]* {{.+}}) |
301 | // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 1, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*, [[SST_TY]]*)* [[OMP_OUTLINED_SST:@.+]] to void |
302 | |
303 | // CHECK: define{{.+}} [[OMP_OUTLINED_SST]](i{{[0-9]+}}* {{.+}}, i{{[0-9]+}}* noalias %{{.+}}, [[SST_TY]]* {{.+}}) |
304 | // CHECK: [[A_PRIV_1:%.+]] = alloca i{{[0-9]+}}, |
305 | // CHECK: store i{{[0-9]+}}* [[A_PRIV_1]], i{{[0-9]+}}** [[A_REF_1:%.+]], |
306 | // CHECK: [[A_REF_VAL_1:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[A_REF_1]] |
307 | // CHECK: [[A_VAL_1:%.+]] = load i{{[0-9]+}}, i{{[0-9]+}}* [[A_REF_VAL_1]] |
308 | // CHECK: [[A_INC_1:%.+]] = add{{.*}} i{{[0-9]+}} [[A_VAL_1]], 1 |
309 | // CHECK: store i{{[0-9]+}} [[A_INC_1]], i{{[0-9]+}}* [[A_REF_VAL_1]], |
310 | // CHECK: ret |
311 | |
312 | #endif |
313 | |
314 | |