1 // expected-no-diagnostics
2 #ifndef HEADER
3 #define HEADER
4 
5 ///==========================================================================///
6 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s --check-prefix CK22 --check-prefix CK22-64
7 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
8 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-64
9 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-32
10 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
11 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-32
12 
13 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp -fopenmp-version=45 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s --check-prefix CK22 --check-prefix CK22-64
14 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp -fopenmp-version=45 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
15 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-version=45 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-64
16 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp -fopenmp-version=45 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-32
17 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp -fopenmp-version=45 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
18 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-version=45 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-32
19 
20 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s --check-prefix CK22 --check-prefix CK22-64
21 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
22 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-64
23 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-32
24 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
25 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  %s  --check-prefix CK22 --check-prefix CK22-32
26 
27 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY21 %s
28 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
29 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY21 %s
30 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -verify -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm %s -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY21 %s
31 // RUN: %clang_cc1 -no-opaque-pointers -DCK22 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s
32 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck -allow-deprecated-dag-overlap  --check-prefix SIMD-ONLY21 %s
33 // SIMD-ONLY21-NOT: {{__kmpc|__tgt}}
34 #ifdef CK22
35 
36 // CK22-DAG: [[ST:%.+]] = type { float }
37 // CK22-DAG: [[STT:%.+]] = type { i32 }
38 
39 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
40 // CK22: [[SIZE00:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
41 // CK22: [[MTYPE00:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
42 
43 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
44 // CK22: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] [i64 400]
45 // CK22: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
46 
47 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
48 // CK22: [[SIZE02:@.+]] = private {{.*}}constant [1 x i64] [i64 {{8|4}}]
49 // CK22: [[MTYPE02:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
50 
51 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
52 // CK22: [[SIZE03:@.+]] = private {{.*}}constant [1 x i64] [i64 16]
53 // CK22: [[MTYPE03:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
54 
55 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
56 // CK22: [[SIZE04:@.+]] = private {{.*}}constant [1 x i64] [i64 20]
57 // CK22: [[MTYPE04:@.+]] = private {{.*}}constant [1 x i64] [i64 51]
58 
59 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
60 // CK22: [[SIZE05:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
61 // CK22: [[MTYPE05:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
62 
63 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
64 // CK22: [[SIZE06:@.+]] = private {{.*}}constant [1 x i64] [i64 400]
65 // CK22: [[MTYPE06:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
66 
67 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
68 // CK22: [[SIZE07:@.+]] = private {{.*}}constant [1 x i64] [i64 {{8|4}}]
69 // CK22: [[MTYPE07:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
70 
71 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
72 // CK22: [[SIZE08:@.+]] = private {{.*}}constant [1 x i64] [i64 16]
73 // CK22: [[MTYPE08:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
74 
75 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
76 // CK22: [[SIZE09:@.+]] = private {{.*}}constant [1 x i64] [i64 20]
77 // CK22: [[MTYPE09:@.+]] = private {{.*}}constant [1 x i64] [i64 51]
78 
79 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
80 // CK22: [[SIZE10:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
81 // CK22: [[MTYPE10:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
82 
83 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
84 // CK22: [[SIZE11:@.+]] = private {{.*}}constant [1 x i64] [i64 400]
85 // CK22: [[MTYPE11:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
86 
87 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
88 // CK22: [[SIZE12:@.+]] = private {{.*}}constant [1 x i64] [i64 {{8|4}}]
89 // CK22: [[MTYPE12:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
90 
91 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
92 // CK22: [[SIZE13:@.+]] = private {{.*}}constant [1 x i64] [i64 16]
93 // CK22: [[MTYPE13:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
94 
95 // CK22-LABEL: @.__omp_offloading_{{.*}}explicit_maps_globals{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
96 // CK22: [[SIZE14:@.+]] = private {{.*}}constant [1 x i64] [i64 20]
97 // CK22: [[MTYPE14:@.+]] = private {{.*}}constant [1 x i64] [i64 51]
98 
99 int a;
100 int c[100];
101 int *d;
102 
103 struct ST {
104   float fa;
105 };
106 
107 ST sa ;
108 ST sc[100];
109 ST *sd;
110 
111 template<typename T>
112 struct STT {
113   T fa;
114 };
115 
116 STT<int> sta ;
117 STT<int> stc[100];
118 STT<int> *std;
119 
120 // CK22-LABEL: explicit_maps_globals{{.*}}(
explicit_maps_globals(void)121 int explicit_maps_globals(void){
122 // Region 00
123 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
124 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
125 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
126 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
127 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
128 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
129 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
130 
131 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
132 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
133 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to i32**
134 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
135 // CK22-DAG: store i32* @a, i32** [[CBP0]]
136 // CK22-DAG: store i32* @a, i32** [[CP0]]
137 
138 // CK22: call void [[CALL00:@.+]](i32* {{[^,]+}})
139 #pragma omp target map(a)
140   { a+=1; }
141 
142 // Region 01
143 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
144 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
145 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
146 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
147 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
148 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
149 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
150 
151 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
152 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
153 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [100 x i32]**
154 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [100 x i32]**
155 // CK22-DAG: store [100 x i32]* @c, [100 x i32]** [[CBP0]]
156 // CK22-DAG: store [100 x i32]* @c, [100 x i32]** [[CP0]]
157 
158 // CK22: call void [[CALL01:@.+]]([100 x i32]* {{[^,]+}})
159 #pragma omp target map(c)
160   { c[3]+=1; }
161 
162 // Region 02
163 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
164 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
165 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
166 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
167 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
168 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
169 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
170 
171 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
172 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
173 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to i32***
174 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32***
175 // CK22-DAG: store i32** @d, i32*** [[CBP0]]
176 // CK22-DAG: store i32** @d, i32*** [[CP0]]
177 
178 // CK22: call void [[CALL02:@.+]](i32** {{[^,]+}})
179 #pragma omp target map(d)
180   { d[3]+=1; }
181 
182 // Region 03
183 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
184 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
185 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
186 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
187 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
188 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
189 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
190 
191 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
192 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
193 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [100 x i32]**
194 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
195 // CK22-DAG: store [100 x i32]* @c, [100 x i32]** [[CBP0]]
196 // CK22-DAG: store i32* getelementptr inbounds ([100 x i32], [100 x i32]* @c, i{{.+}} 0, i{{.+}} 1), i32** [[CP0]]
197 
198 // CK22: call void [[CALL03:@.+]]([100 x i32]* {{[^,]+}})
199 #pragma omp target map(c [1:4])
200   { c[3]+=1; }
201 
202 // Region 04
203 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
204 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
205 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
206 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
207 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
208 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
209 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
210 
211 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
212 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
213 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to i32***
214 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
215 // CK22-DAG: store i32** @d, i32*** [[CBP0]]
216 // CK22-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
217 // CK22-DAG: [[SEC0]] = getelementptr {{.*}}i32* [[RVAR00:%.+]], i{{.+}} 2
218 // CK22-DAG: [[RVAR00]] = load i32*, i32** @d
219 
220 // CK22: call void [[CALL04:@.+]](i32* {{[^,]+}})
221 #pragma omp target map(d [2:5])
222   { d[3]+=1; }
223 
224 // Region 05
225 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
226 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
227 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
228 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
229 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
230 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
231 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
232 
233 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
234 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
235 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[ST]]**
236 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[ST]]**
237 // CK22-DAG: store [[ST]]* @sa, [[ST]]** [[CBP0]]
238 // CK22-DAG: store [[ST]]* @sa, [[ST]]** [[CP0]]
239 
240 // CK22: call void [[CALL05:@.+]]([[ST]]* {{[^,]+}})
241 #pragma omp target map(sa)
242   { sa.fa+=1; }
243 
244 // Region 06
245 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
246 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
247 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
248 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
249 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
250 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
251 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
252 
253 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
254 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
255 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [100 x [[ST]]]**
256 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [100 x [[ST]]]**
257 // CK22-DAG: store [100 x [[ST]]]* @sc, [100 x [[ST]]]** [[CBP0]]
258 // CK22-DAG: store [100 x [[ST]]]* @sc, [100 x [[ST]]]** [[CP0]]
259 
260 // CK22: call void [[CALL06:@.+]]([100 x [[ST]]]* {{[^,]+}})
261 #pragma omp target map(sc)
262   { sc[3].fa+=1; }
263 
264 // Region 07
265 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
266 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
267 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
268 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
269 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
270 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
271 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
272 
273 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
274 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
275 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[ST]]***
276 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[ST]]***
277 // CK22-DAG: store [[ST]]** @sd, [[ST]]*** [[CBP0]]
278 // CK22-DAG: store [[ST]]** @sd, [[ST]]*** [[CP0]]
279 
280 // CK22: call void [[CALL07:@.+]]([[ST]]** {{[^,]+}})
281 #pragma omp target map(sd)
282   { sd[3].fa+=1; }
283 
284 // Region 08
285 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
286 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
287 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
288 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
289 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
290 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
291 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
292 
293 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
294 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
295 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [100 x [[ST]]]**
296 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[ST]]**
297 // CK22-DAG: store [100 x [[ST]]]* @sc, [100 x [[ST]]]** [[CBP0]]
298 // CK22-DAG: store [[ST]]* getelementptr inbounds ([100 x [[ST]]], [100 x [[ST]]]* @sc, i{{.+}} 0, i{{.+}} 1), [[ST]]** [[CP0]]
299 
300 // CK22: call void [[CALL08:@.+]]([100 x [[ST]]]* {{[^,]+}})
301 #pragma omp target map(sc [1:4])
302   { sc[3].fa+=1; }
303 
304 // Region 09
305 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
306 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
307 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
308 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
309 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
310 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
311 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
312 
313 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
314 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
315 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[ST]]***
316 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[ST]]**
317 // CK22-DAG: store [[ST]]** @sd, [[ST]]*** [[CBP0]]
318 // CK22-DAG: store [[ST]]* [[SEC0:%.+]], [[ST]]** [[CP0]]
319 // CK22-DAG: [[SEC0]] = getelementptr {{.*}}[[ST]]* [[RVAR00:%.+]], i{{.+}} 2
320 // CK22-DAG: [[RVAR00]] = load [[ST]]*, [[ST]]** @sd
321 
322 // CK22: call void [[CALL09:@.+]]([[ST]]* {{[^,]+}})
323 #pragma omp target map(sd [2:5])
324   { sd[3].fa+=1; }
325 
326 // Region 10
327 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
328 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
329 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
330 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
331 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
332 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
333 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
334 
335 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
336 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
337 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[STT]]**
338 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[STT]]**
339 // CK22-DAG: store [[STT]]* @sta, [[STT]]** [[CBP0]]
340 // CK22-DAG: store [[STT]]* @sta, [[STT]]** [[CP0]]
341 
342 // CK22: call void [[CALL10:@.+]]([[STT]]* {{[^,]+}})
343 #pragma omp target map(sta)
344   { sta.fa+=1; }
345 
346 // Region 11
347 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
348 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
349 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
350 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
351 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
352 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
353 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
354 
355 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
356 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
357 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [100 x [[STT]]]**
358 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [100 x [[STT]]]**
359 // CK22-DAG: store [100 x [[STT]]]* @stc, [100 x [[STT]]]** [[CBP0]]
360 // CK22-DAG: store [100 x [[STT]]]* @stc, [100 x [[STT]]]** [[CP0]]
361 
362 // CK22: call void [[CALL11:@.+]]([100 x [[STT]]]* {{[^,]+}})
363 #pragma omp target map(stc)
364   { stc[3].fa+=1; }
365 
366 // Region 12
367 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
368 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
369 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
370 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
371 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
372 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
373 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
374 
375 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
376 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
377 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[STT]]***
378 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[STT]]***
379 // CK22-DAG: store [[STT]]** @std, [[STT]]*** [[CBP0]]
380 // CK22-DAG: store [[STT]]** @std, [[STT]]*** [[CP0]]
381 
382 // CK22: call void [[CALL12:@.+]]([[STT]]** {{[^,]+}})
383 #pragma omp target map(std)
384   { std[3].fa+=1; }
385 
386 // Region 13
387 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
388 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
389 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
390 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
391 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
392 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
393 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
394 
395 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
396 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
397 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [100 x [[STT]]]**
398 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[STT]]**
399 // CK22-DAG: store [100 x [[STT]]]* @stc, [100 x [[STT]]]** [[CBP0]]
400 // CK22-DAG: store [[STT]]* getelementptr inbounds ([100 x [[STT]]], [100 x [[STT]]]* @stc, i{{.+}} 0, i{{.+}} 1),  [[STT]]** [[CP0]]
401 
402 // CK22: call void [[CALL13:@.+]]([100 x [[STT]]]* {{[^,]+}})
403 #pragma omp target map(stc [1:4])
404   { stc[3].fa+=1; }
405 
406 // Region 14
407 // CK22-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
408 // CK22-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
409 // CK22-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
410 // CK22-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
411 // CK22-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
412 // CK22-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
413 // CK22-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
414 
415 // CK22-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
416 // CK22-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
417 // CK22-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[STT]]***
418 // CK22-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[STT]]**
419 // CK22-DAG: store [[STT]]** @std, [[STT]]*** [[CBP0]]
420 // CK22-DAG: store [[STT]]* [[SEC0:%.+]], [[STT]]** [[CP0]]
421 // CK22-DAG: [[SEC0]] = getelementptr {{.*}}[[STT]]* [[RVAR00:%.+]], i{{.+}} 2
422 // CK22-DAG: [[RVAR00]] = load [[STT]]*, [[STT]]** @std
423 
424 // CK22: call void [[CALL14:@.+]]([[STT]]* {{[^,]+}})
425 #pragma omp target map(std [2:5])
426   { std[3].fa+=1; }
427 
428   return 0;
429 }
430 // CK22: define {{.+}}[[CALL00]]
431 // CK22: define {{.+}}[[CALL01]]
432 // CK22: define {{.+}}[[CALL02]]
433 // CK22: define {{.+}}[[CALL03]]
434 // CK22: define {{.+}}[[CALL04]]
435 // CK22: define {{.+}}[[CALL05]]
436 // CK22: define {{.+}}[[CALL06]]
437 // CK22: define {{.+}}[[CALL07]]
438 // CK22: define {{.+}}[[CALL08]]
439 // CK22: define {{.+}}[[CALL09]]
440 // CK22: define {{.+}}[[CALL10]]
441 // CK22: define {{.+}}[[CALL11]]
442 // CK22: define {{.+}}[[CALL12]]
443 // CK22: define {{.+}}[[CALL13]]
444 // CK22: define {{.+}}[[CALL14]]
445 #endif // CK22
446 #endif
447