1 // expected-no-diagnostics
2 #ifndef HEADER
3 #define HEADER
4 
5 ///==========================================================================///
6 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-64
7 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-64
9 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-32
10 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-32
12 
13 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-64
14 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-64
16 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-32
17 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-32
19 
20 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-64
21 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-64
23 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-32
24 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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 CK24 --check-prefix CK24-32
26 
27 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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-ONLY23 %s
28 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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-ONLY23 %s
30 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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-ONLY23 %s
31 // RUN: %clang_cc1 -no-opaque-pointers -DCK24 -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-ONLY23 %s
33 // SIMD-ONLY23-NOT: {{__kmpc|__tgt}}
34 #ifdef CK24
35 
36 // CK24-DAG: [[SC:%.+]] = type { i32, [[SB:%.+]], [[SB:%.+]]*, [10 x i32] }
37 // CK24-DAG: [[SB]] = type { i32, [[SA:%.+]], [10 x [[SA:%.+]]], [10 x [[SA:%.+]]*], [[SA:%.+]]* }
38 // CK24-DAG: [[SA]] = type { i32, [[SA]]*, [10 x i32] }
39 
40 struct SA{
41   int a;
42   struct SA *p;
43   int b[10];
44 };
45 struct SB{
46   int a;
47   struct SA s;
48   struct SA sa[10];
49   struct SA *sp[10];
50   struct SA *p;
51 };
52 struct SC{
53   int a;
54   struct SB s;
55   struct SB *p;
56   int b[10];
57 };
58 
59 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
60 // CK24: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
61 // CK24: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
62 
63 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
64 // CK24: [[SIZE13:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
65 // CK24: [[MTYPE13:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
66 
67 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
68 // CK24: [[SIZE14:@.+]] = private {{.*}}constant [1 x i64] [i64 {{48|56}}]
69 // CK24: [[MTYPE14:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
70 
71 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
72 // CK24: [[SIZE15:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
73 // CK24: [[MTYPE15:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
74 
75 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
76 // CK24: [[SIZE16:@.+]] = private {{.*}}constant [2 x i64] [i64 0, i64 20]
77 // CK24: [[MTYPE16:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659]
78 
79 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
80 // CK24: [[SIZE17:@.+]] = private {{.*}}constant [2 x i64] [i64 0, i64 {{3560|2880}}]
81 // CK24: [[MTYPE17:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710675]
82 
83 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
84 // CK24: [[SIZE18:@.+]] = private {{.*}}constant [1 x i64] [i64 4]
85 // CK24: [[MTYPE18:@.+]] = private {{.*}}constant [1 x i64] [i64 35]
86 
87 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
88 // CK24: [[SIZE19:@.+]] = private {{.*}}constant [3 x i64] [i64 0, i64 {{8|4}}, i64 4]
89 // CK24: [[MTYPE19:@.+]] = private {{.*}}constant [3 x i64] [i64 32, i64 281474976710659, i64 281474976710675]
90 
91 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
92 // CK24: [[SIZE20:@.+]] = private {{.*}}constant [2 x i64] [i64 0, i64 4]
93 // CK24: [[MTYPE20:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710675]
94 
95 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
96 // CK24: [[SIZE21:@.+]] = private {{.*}}constant [3 x i64] [i64 0, i64 {{8|4}}, i64 4]
97 // CK24: [[MTYPE21:@.+]] = private {{.*}}constant [3 x i64] [i64 32, i64 281474976710659, i64 281474976710675]
98 
99 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
100 // CK24: [[SIZE22:@.+]] = private {{.*}}constant [2 x i64] [i64 0, i64 8]
101 // CK24: [[MTYPE22:@.+]] = private {{.*}}constant [2 x i64] [i64 32, i64 281474976710659]
102 
103 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
104 // CK24: [[SIZE23:@.+]] = private {{.*}}constant [3 x i64] [i64 0, i64 {{8|4}}, i64 8]
105 // CK24: [[MTYPE23:@.+]] = private {{.*}}constant [3 x i64] [i64 32, i64 281474976710659, i64 281474976710675]
106 
107 // CK24-LABEL: @.__omp_offloading_{{.*}}explicit_maps_struct_fields{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0
108 // CK24: [[SIZE24:@.+]] = private {{.*}}constant [4 x i64] [i64 0, i64 {{8|4}}, i64 {{8|4}}, i64 4]
109 // CK24: [[MTYPE24:@.+]] = private {{.*}}constant [4 x i64] [i64 32, i64 281474976710672, i64 16, i64 19]
110 
111 // CK24-LABEL: explicit_maps_struct_fields
explicit_maps_struct_fields(int a)112 int explicit_maps_struct_fields(int a){
113   SC s;
114   SC *p;
115 
116 // Region 01
117 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
118 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
119 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
120 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
121 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
122 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
123 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
124 
125 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
126 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
127 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
128 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
129 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
130 // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
131 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR0]], i{{.+}} 0, i{{.+}} 0
132 
133 // CK24: call void [[CALL01:@.+]]([[SC]]* {{[^,]+}})
134 #pragma omp target map(s.a)
135   { s.a++; }
136 
137 //
138 // Same thing but starting from a pointer.
139 //
140 // Region 13
141 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
142 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
143 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
144 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
145 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
146 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
147 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
148 
149 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
150 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
151 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
152 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
153 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
154 // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
155 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 0
156 
157 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
158 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
159 
160 // CK24: call void [[CALL13:@.+]]([[SC]]* {{[^,]+}})
161 #pragma omp target map(p->a)
162   { p->a++; }
163 
164 // Region 14
165 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
166 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
167 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
168 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
169 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
170 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
171 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
172 
173 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
174 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
175 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
176 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SA]]**
177 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
178 // CK24-DAG: store [[SA]]* [[SEC0:%.+]], [[SA]]** [[CP0]]
179 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SB]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 1
180 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
181 
182 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
183 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
184 
185 // CK24: call void [[CALL14:@.+]]([[SC]]* {{[^,]+}})
186 #pragma omp target map(p->s.s)
187   { p->a++; }
188 
189 // Region 15
190 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
191 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
192 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
193 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
194 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
195 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
196 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
197 
198 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
199 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
200 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
201 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
202 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
203 // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
204 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SA]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0
205 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SB]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 1
206 // CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
207 
208 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
209 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
210 
211 // CK24: call void [[CALL15:@.+]]([[SC]]* {{[^,]+}})
212 #pragma omp target map(p->s.s.a)
213   { p->a++; }
214 
215 // Region 16
216 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
217 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
218 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
219 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
220 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
221 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
222 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
223 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
224 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
225 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
226 
227 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
228 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
229 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
230 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
231 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
232 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
233 // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
234 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
235 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[10 x i32]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0
236 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 3
237 
238 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
239 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
240 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
241 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
242 // CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
243 // CK24-DAG: store i32* [[SEC0]], i32** [[CP1]]
244 
245 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
246 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
247 
248 // CK24: call void [[CALL16:@.+]]([[SC]]* {{[^,]+}})
249 #pragma omp target map(p->b[:5])
250   { p->a++; }
251 
252 // Region 17
253 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
254 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
255 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
256 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
257 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
258 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
259 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
260 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
261 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
262 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
263 
264 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
265 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
266 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
267 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
268 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SB]]***
269 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
270 // CK24-DAG: store [[SB]]** [[SEC0:%.+]], [[SB]]*** [[CP0]]
271 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
272 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 2
273 
274 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
275 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
276 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SB]]***
277 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to [[SB]]**
278 // CK24-DAG: store [[SB]]** [[SEC0]], [[SB]]*** [[CBP1]]
279 // CK24-DAG: store [[SB]]* [[SEC1:%.+]], [[SB]]** [[CP1]]
280 // CK24-DAG: [[SEC1]] = getelementptr {{.*}}[[SB]]* [[SEC11:%[^,]+]], i{{.+}} 0
281 // CK24-DAG: [[SEC11]] = load [[SB]]*, [[SB]]** [[SEC111:%[^,]+]],
282 // CK24-DAG: [[SEC111]] = getelementptr {{.*}}[[SC]]* [[VAR000:%.+]], i{{.+}} 0, i{{.+}} 2
283 
284 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
285 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
286 // CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}}
287 
288 // CK24: call void [[CALL17:@.+]]([[SC]]* {{[^,]+}})
289 #pragma omp target map(p->p[:5])
290   { p->a++; }
291 
292 // Region 18
293 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
294 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
295 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
296 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
297 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
298 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
299 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
300 
301 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
302 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
303 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
304 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
305 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
306 // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
307 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SA]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0
308 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[10 x [[SA]]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 3
309 // CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SB]]* [[SEC0000:%[^,]+]], i{{.+}} 0, i{{.+}} 2
310 // CK24-DAG: [[SEC0000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
311 
312 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
313 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
314 
315 // CK24: call void [[CALL18:@.+]]([[SC]]* {{[^,]+}})
316 #pragma omp target map(p->s.sa[3].a)
317   { p->a++; }
318 
319 // Region 19
320 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
321 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
322 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
323 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
324 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
325 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
326 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
327 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
328 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
329 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
330 
331 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
332 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
333 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
334 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
335 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SA]]***
336 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
337 // CK24-DAG: store [[SA]]** [[SEC0:%.+]], [[SA]]*** [[CP0]]
338 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
339 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[10 x [[SA]]*]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 3
340 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SB]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 3
341 // CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
342 
343 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
344 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
345 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
346 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to [[SA]]***
347 // CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
348 // CK24-DAG: store [[SA]]** [[SEC0]], [[SA]]*** [[CP1]]
349 
350 // CK24-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 2
351 // CK24-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 2
352 // CK24-DAG: [[CBP2:%.+]] = bitcast i8** [[BP2]] to [[SA]]***
353 // CK24-DAG: [[CP2:%.+]] = bitcast i8** [[P2]] to i32**
354 // CK24-DAG: store [[SA]]** [[SEC0]], [[SA]]*** [[CBP2]]
355 // CK24-DAG: store i32* [[SEC1:%.+]], i32** [[CP2]]
356 // CK24-DAG: [[SEC1]] = getelementptr {{.*}}[[SA]]* [[SEC11:%[^,]+]], i{{.+}} 0, i{{.+}} 0
357 // CK24-DAG: [[SEC11]] = load [[SA]]*, [[SA]]** [[SEC111:%[^,]+]],
358 // CK24-DAG: [[SEC111]] = getelementptr {{.*}}[10 x [[SA]]*]* [[SEC1111:%[^,]+]], i{{.+}} 0, i{{.+}} 3
359 // CK24-DAG: [[SEC1111]] = getelementptr {{.*}}[[SB]]* [[SEC11111:%[^,]+]], i{{.+}} 0, i{{.+}} 3
360 // CK24-DAG: [[SEC11111]] = getelementptr {{.*}}[[SC]]* [[VAR000:%.+]], i{{.+}} 0, i{{.+}} 1
361 
362 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
363 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
364 // CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}}
365 
366 // CK24: call void [[CALL19:@.+]]([[SC]]* {{[^,]+}})
367 #pragma omp target map(p->s.sp[3]->a)
368   { p->a++; }
369 
370 // Region 20
371 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
372 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
373 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
374 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
375 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
376 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
377 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
378 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
379 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
380 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
381 
382 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
383 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
384 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
385 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
386 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SB]]***
387 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
388 // CK24-DAG: store [[SB]]** [[SEC0:%.+]], [[SB]]*** [[CP0]]
389 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
390 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 2
391 
392 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
393 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
394 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SB]]***
395 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
396 // CK24-DAG: store [[SB]]** [[SEC0]], [[SB]]*** [[CBP1]]
397 // CK24-DAG: store i32* [[SEC1:%.+]], i32** [[CP1]]
398 // CK24-DAG: [[SEC1]] = getelementptr {{.*}}[[SB]]* [[SEC11:%[^,]+]], i{{.+}} 0
399 // CK24-DAG: [[SEC11]] = load [[SB]]*, [[SB]]** [[SEC111:%[^,]+]],
400 // CK24-DAG: [[SEC111]] = getelementptr {{.*}}[[SC]]* [[VAR000:%.+]], i{{.+}} 0, i{{.+}} 2
401 
402 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
403 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
404 // CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}}
405 
406 // CK24: call void [[CALL20:@.+]]([[SC]]* {{[^,]+}})
407 #pragma omp target map(p->p->a)
408   { p->a++; }
409 
410 // Region 21
411 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
412 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
413 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
414 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
415 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
416 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
417 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
418 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
419 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
420 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
421 
422 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
423 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
424 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
425 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
426 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SA]]***
427 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
428 // CK24-DAG: store [[SA]]** [[SEC0:%.+]], [[SA]]*** [[CP0]]
429 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
430 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SB]]* [[SEC00:[^,]+]], i{{.+}} 0, i{{.+}} 4
431 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
432 
433 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
434 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
435 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
436 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to [[SA]]***
437 // CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
438 // CK24-DAG: store [[SA]]** [[SEC0]], [[SA]]*** [[CP1]]
439 
440 // CK24-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 2
441 // CK24-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 2
442 // CK24-DAG: [[CBP2:%.+]] = bitcast i8** [[BP2]] to [[SA]]***
443 // CK24-DAG: [[CP2:%.+]] = bitcast i8** [[P2]] to i32**
444 // CK24-DAG: store [[SA]]** [[SEC0]], [[SA]]*** [[CBP2]]
445 // CK24-DAG: store i32* [[SEC1:%.+]], i32** [[CP2]]
446 // CK24-DAG: [[SEC1]] = getelementptr {{.*}}[[SA]]* [[SEC11:%[^,]+]], i{{.+}} 0
447 // CK24-DAG: [[SEC11]] = load [[SA]]*, [[SA]]** [[SEC111:%[^,]+]],
448 // CK24-DAG: [[SEC111]] = getelementptr {{.*}}[[SB]]* [[SEC1111:[^,]+]], i{{.+}} 0, i{{.+}} 4
449 // CK24-DAG: [[SEC1111]] = getelementptr {{.*}}[[SC]]* [[VAR000:%.+]], i{{.+}} 0, i{{.+}} 1
450 
451 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
452 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
453 // CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}}
454 
455 // CK24: call void [[CALL21:@.+]]([[SC]]* {{[^,]+}})
456 #pragma omp target map(p->s.p->a)
457   { p->a++; }
458 
459 // Region 22
460 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
461 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
462 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
463 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
464 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
465 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
466 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
467 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
468 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
469 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
470 
471 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
472 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
473 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
474 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
475 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32**
476 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
477 // CK24-DAG: store i32* [[SEC0:%.+]], i32** [[CP0]]
478 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
479 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[10 x i32]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 0
480 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SA]]* [[SEC000:%[^,]+]], i{{.+}} 0, i{{.+}} 2
481 // CK24-DAG: [[SEC000]] = getelementptr {{.*}}[[SB]]* [[SEC0000:%[^,]+]], i{{.+}} 0, i{{.+}} 1
482 // CK24-DAG: [[SEC0000]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
483 
484 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
485 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
486 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
487 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to i32**
488 // CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
489 // CK24-DAG: store i32* [[SEC0]], i32** [[CP1]]
490 
491 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
492 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
493 
494 // CK24: call void [[CALL22:@.+]]([[SC]]* {{[^,]+}})
495 #pragma omp target map(p->s.s.b[:2])
496   { p->a++; }
497 
498 // Region 23
499 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
500 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
501 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
502 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
503 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
504 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
505 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
506 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
507 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
508 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
509 
510 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
511 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
512 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
513 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
514 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SA]]***
515 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
516 // CK24-DAG: store [[SA]]** [[SEC0:%.+]], [[SA]]*** [[CP0]]
517 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
518 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SB]]* [[SEC00:%[^,]+]], i{{.+}} 0, i{{.+}} 4
519 // CK24-DAG: [[SEC00]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 1
520 
521 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
522 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
523 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SC]]**
524 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to [[SA]]***
525 // CK24-DAG: store [[SC]]* [[VAR0]], [[SC]]** [[CBP1]]
526 // CK24-DAG: store [[SA]]** [[SEC0]], [[SA]]*** [[CP1]]
527 
528 // CK24-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 2
529 // CK24-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 2
530 // CK24-DAG: [[CBP2:%.+]] = bitcast i8** [[BP2]] to [[SA]]***
531 // CK24-DAG: [[CP2:%.+]] = bitcast i8** [[P2]] to i32**
532 // CK24-DAG: store [[SA]]** [[SEC0]], [[SA]]*** [[CBP2]]
533 // CK24-DAG: store i32* [[SEC1:%.+]], i32** [[CP2]]
534 // CK24-DAG: [[SEC1]] = getelementptr {{.*}}[10 x i32]* [[SEC11:%[^,]+]], i{{.+}} 0, i{{.+}} 0
535 // CK24-DAG: [[SEC11]] = getelementptr {{.*}}[[SA]]* [[SEC111:%[^,]+]], i{{.+}} 0, i{{.+}} 2
536 // CK24-DAG: [[SEC111]] = load [[SA]]*, [[SA]]** [[SEC1111:%[^,]+]],
537 // CK24-DAG: [[SEC1111]] = getelementptr {{.*}}[[SB]]* [[SEC11111:%[^,]+]], i{{.+}} 0, i{{.+}} 4
538 // CK24-DAG: [[SEC11111]] = getelementptr {{.*}}[[SC]]* [[VAR000:%.+]], i{{.+}} 0, i{{.+}} 1
539 
540 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
541 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
542 // CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}}
543 
544 // CK24: call void [[CALL23:@.+]]([[SC]]* {{[^,]+}})
545 #pragma omp target map(p->s.p->b[:2])
546   { p->a++; }
547 
548 // Region 24
549 // CK24-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]])
550 // CK24-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2
551 // CK24-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]]
552 // CK24-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3
553 // CK24-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]]
554 // CK24-DAG: [[SARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 4
555 // CK24-DAG: store i64* [[SIZES:%.+]], i64** [[SARG]]
556 // CK24-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]
557 // CK24-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]
558 // CK24-DAG: [[SIZES]] = getelementptr inbounds {{.+}}[[S:%[^,]+]]
559 
560 // CK24-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0
561 // CK24-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0
562 // CK24-DAG: [[S0:%.+]] = getelementptr inbounds {{.+}}[[S]], i{{.+}} 0, i{{.+}} 0
563 // CK24-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to [[SC]]**
564 // CK24-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to [[SB]]***
565 // CK24-DAG: store [[SC]]* [[VAR0:%.+]], [[SC]]** [[CBP0]]
566 // CK24-DAG: store [[SB]]** [[SEC0:%.+]], [[SB]]*** [[CP0]]
567 // CK24-DAG: store i64 {{%.+}}, i64* [[S0]]
568 // CK24-DAG: [[SEC0]] = getelementptr {{.*}}[[SC]]* [[VAR00:%.+]], i{{.+}} 0, i{{.+}} 2
569 
570 // CK24-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 1
571 // CK24-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 1
572 // CK24-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to [[SB]]***
573 // CK24-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to [[SA]]***
574 // CK24-DAG: store [[SB]]** [[SEC0]], [[SB]]*** [[CBP1]]
575 // CK24-DAG: store [[SA]]** [[SEC1:%.+]], [[SA]]*** [[CP1]]
576 // CK24-DAG: [[SEC1]] = getelementptr {{.*}}[[SB]]* [[SEC11:%[^,]+]], i{{.+}} 0, i{{.+}} 4
577 // CK24-DAG: [[SEC11]] = load [[SB]]*, [[SB]]** [[SEC111:%[^,]+]],
578 // CK24-DAG: [[SEC111]] = getelementptr {{.*}}[[SC]]* [[VAR000:%.+]], i{{.+}} 0, i{{.+}} 2
579 
580 // CK24-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 2
581 // CK24-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 2
582 // CK24-DAG: [[CBP2:%.+]] = bitcast i8** [[BP2]] to [[SA]]***
583 // CK24-DAG: [[CP2:%.+]] = bitcast i8** [[P2]] to [[SA]]***
584 // CK24-DAG: store [[SA]]** [[SEC1]], [[SA]]*** [[CBP2]]
585 // CK24-DAG: store [[SA]]** [[SEC2:%.+]], [[SA]]*** [[CP2]]
586 // CK24-DAG: [[SEC2]] = getelementptr {{.*}}[[SA]]* [[SEC22:%[^,]+]], i{{.+}} 0, i{{.+}} 1
587 // CK24-DAG: [[SEC22]] = load [[SA]]*, [[SA]]** [[SEC222:%[^,]+]],
588 // CK24-DAG: [[SEC222]] = getelementptr {{.*}}[[SB]]* [[SEC2222:%[^,]+]], i{{.+}} 0, i{{.+}} 4
589 // CK24-DAG: [[SEC2222]] = load [[SB]]*, [[SB]]** [[SEC22222:%[^,]+]],
590 // CK24-DAG: [[SEC22222]] = getelementptr {{.*}}[[SC]]* [[VAR0000:%.+]], i{{.+}} 0, i{{.+}} 2
591 
592 // CK24-DAG: [[BP3:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 3
593 // CK24-DAG: [[P3:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 3
594 // CK24-DAG: [[CBP3:%.+]] = bitcast i8** [[BP3]] to [[SA]]***
595 // CK24-DAG: [[CP3:%.+]] = bitcast i8** [[P3]] to i32**
596 // CK24-DAG: store [[SA]]** [[SEC2]], [[SA]]*** [[CBP3]]
597 // CK24-DAG: store i32* [[SEC3:%.+]], i32** [[CP3]]
598 // CK24-DAG: [[SEC3]] = getelementptr {{.*}}[[SA]]* [[SEC33:%[^,]+]], i{{.+}} 0, i{{.+}} 0
599 // CK24-DAG: [[SEC33]] = load [[SA]]*, [[SA]]** [[SEC333:%[^,]+]],
600 // CK24-DAG: [[SEC333]] = getelementptr {{.*}}[[SA]]* [[SEC3333:%[^,]+]], i{{.+}} 0, i{{.+}} 1
601 // CK24-DAG: [[SEC3333]] = load [[SA]]*, [[SA]]** [[SEC33333:%[^,]+]],
602 // CK24-DAG: [[SEC33333]] = getelementptr {{.*}}[[SB]]* [[SEC333333:%[^,]+]], i{{.+}} 0, i{{.+}} 4
603 // CK24-DAG: [[SEC333333]] = load [[SB]]*, [[SB]]** [[SEC3333333:%[^,]+]],
604 // CK24-DAG: [[SEC3333333]] = getelementptr {{.*}}[[SC]]* [[VAR00000:%.+]], i{{.+}} 0, i{{.+}} 2
605 
606 // CK24-DAG: [[VAR0]] = load [[SC]]*, [[SC]]** %{{.+}}
607 // CK24-DAG: [[VAR00]] = load [[SC]]*, [[SC]]** %{{.+}}
608 // CK24-DAG: [[VAR000]] = load [[SC]]*, [[SC]]** %{{.+}}
609 // CK24-DAG: [[VAR0000]] = load [[SC]]*, [[SC]]** %{{.+}}
610 // CK24-DAG: [[VAR00000]] = load [[SC]]*, [[SC]]** %{{.+}}
611 
612 // CK24: call void [[CALL24:@.+]]([[SC]]* {{[^,]+}})
613 #pragma omp target map(p->p->p->p->a)
614   { p->a++; }
615 
616   return s.a;
617 }
618 
619 // CK24: define {{.+}}[[CALL01]]
620 // CK24: define {{.+}}[[CALL13]]
621 // CK24: define {{.+}}[[CALL14]]
622 // CK24: define {{.+}}[[CALL15]]
623 // CK24: define {{.+}}[[CALL16]]
624 // CK24: define {{.+}}[[CALL17]]
625 // CK24: define {{.+}}[[CALL18]]
626 // CK24: define {{.+}}[[CALL19]]
627 // CK24: define {{.+}}[[CALL20]]
628 // CK24: define {{.+}}[[CALL21]]
629 // CK24: define {{.+}}[[CALL22]]
630 // CK24: define {{.+}}[[CALL23]]
631 // CK24: define {{.+}}[[CALL24]]
632 #endif // CK24
633 #endif
634