1 // NOTE: Assertions have been autogenerated by utils/update_cc_test_checks.py UTC_ARGS: --check-globals --replace-value-regex "__omp_offloading_[0-9a-z]+_[0-9a-z]+" --prefix-filecheck-ir-name _ --global-value-regex ".offload_maptypes.*" ".offload_sizes.*" --global-hex-value-regex ".offload_maptypes.*"
2 // expected-no-diagnostics
3 #ifndef HEADER
4 #define HEADER
5
6 //--------------------------------------------------
7 // With -DUSE.
8 //--------------------------------------------------
9
10 // powerpc64le-ibm-linux-gnu
11
12 // RUN: %clang_cc1 -no-opaque-pointers -DUSE -verify -fopenmp -fopenmp-extensions \
13 // RUN: -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ \
14 // RUN: -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | \
15 // RUN: FileCheck %s --check-prefixes=CHECK-USE-PPC64LE
16 // RUN: %clang_cc1 -no-opaque-pointers -DUSE -fopenmp -fopenmp-extensions \
17 // RUN: -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 \
18 // RUN: -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
19 // RUN: %clang_cc1 -no-opaque-pointers -DUSE -fopenmp -fopenmp-extensions \
20 // RUN: -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ \
21 // RUN: -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t \
22 // RUN: -verify %s -emit-llvm -o - | \
23 // RUN: FileCheck %s --check-prefixes=CHECK-USE-PPC64LE
24
25 // i386-pc-linux-gnu
26
27 // RUN: %clang_cc1 -no-opaque-pointers -DUSE -verify -fopenmp -fopenmp-extensions \
28 // RUN: -fopenmp-targets=i386-pc-linux-gnu -x c++ \
29 // RUN: -triple i386-unknown-unknown -emit-llvm %s -o - | \
30 // RUN: FileCheck %s --check-prefixes=CHECK-USE-I386
31 // RUN: %clang_cc1 -no-opaque-pointers -DUSE -fopenmp -fopenmp-extensions \
32 // RUN: -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 \
33 // RUN: -triple i386-unknown-unknown -emit-pch -o %t %s
34 // RUN: %clang_cc1 -no-opaque-pointers -DUSE -fopenmp -fopenmp-extensions \
35 // RUN: -fopenmp-targets=i386-pc-linux-gnu -x c++ \
36 // RUN: -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s \
37 // RUN: -emit-llvm -o - | \
38 // RUN: FileCheck %s --check-prefixes=CHECK-USE-I386
39
40 //--------------------------------------------------
41 // Without -DUSE.
42 //--------------------------------------------------
43
44 // powerpc64le-ibm-linux-gnu
45
46 // RUN: %clang_cc1 -no-opaque-pointers -verify -fopenmp -fopenmp-extensions \
47 // RUN: -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ \
48 // RUN: -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | \
49 // RUN: FileCheck %s --check-prefixes=CHECK-NOUSE-PPC64LE
50 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-extensions \
51 // RUN: -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 \
52 // RUN: -triple powerpc64le-unknown-unknown -emit-pch -o %t %s
53 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-extensions \
54 // RUN: -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ \
55 // RUN: -triple powerpc64le-unknown-unknown -std=c++11 -include-pch %t \
56 // RUN: -verify %s -emit-llvm -o - | \
57 // RUN: FileCheck %s --check-prefixes=CHECK-NOUSE-PPC64LE
58
59 // i386-pc-linux-gnu
60
61 // RUN: %clang_cc1 -no-opaque-pointers -verify -fopenmp -fopenmp-extensions \
62 // RUN: -fopenmp-targets=i386-pc-linux-gnu -x c++ \
63 // RUN: -triple i386-unknown-unknown -emit-llvm %s -o - | \
64 // RUN: FileCheck %s --check-prefixes=CHECK-NOUSE-I386
65 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-extensions \
66 // RUN: -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 \
67 // RUN: -triple i386-unknown-unknown -emit-pch -o %t %s
68 // RUN: %clang_cc1 -no-opaque-pointers -fopenmp -fopenmp-extensions \
69 // RUN: -fopenmp-targets=i386-pc-linux-gnu -x c++ \
70 // RUN: -triple i386-unknown-unknown -std=c++11 -include-pch %t -verify %s \
71 // RUN: -emit-llvm -o - | \
72 // RUN: FileCheck %s --check-prefixes=CHECK-NOUSE-I386
73
74 // Map flags used in @.offload_maptypes* below:
75 //
76 // TO = 0x1
77 // FROM = 0x2
78 // ALWAYS = 0x4
79 // TARGET_PARAM = 0x20
80 // CLOSE = 0x400
81 // OMPX_HOLD = 0x2000
82 // MEMBER_OF_1 = 0x1000000000000
83 // MEMBER_OF_5 = 0x5000000000000
84
85 //.
86 // CHECK-USE-PPC64LE: @.offload_sizes = private unnamed_addr constant [7 x i64] [i64 0, i64 4, i64 4, i64 4, i64 0, i64 4, i64 4]
87 // CHECK-USE-PPC64LE: @.offload_maptypes = private unnamed_addr constant [7 x i64] [i64 [[#0x2020]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]], i64 [[#0x2023]], i64 [[#0x2020]], i64 [[#0x5000000002003]], i64 [[#0x5000000002003]]]
88 // CHECK-USE-PPC64LE: @.offload_sizes.1 = private unnamed_addr constant [1 x i64] [i64 4]
89 // CHECK-USE-PPC64LE: @.offload_maptypes.2 = private unnamed_addr constant [1 x i64] [i64 [[#0x2427]]]
90 // CHECK-USE-PPC64LE: @.offload_sizes.3 = private unnamed_addr constant [3 x i64] [i64 0, i64 4, i64 4]
91 // CHECK-USE-PPC64LE: @.offload_maptypes.4 = private unnamed_addr constant [3 x i64] [i64 [[#0x2020]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]]]
92 //.
93 // CHECK-USE-I386: @.offload_sizes = private unnamed_addr constant [7 x i64] [i64 0, i64 4, i64 4, i64 4, i64 0, i64 4, i64 4]
94 // CHECK-USE-I386: @.offload_maptypes = private unnamed_addr constant [7 x i64] [i64 [[#0x2020]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]], i64 [[#0x2023]], i64 [[#0x2020]], i64 [[#0x5000000002003]], i64 [[#0x5000000002003]]]
95 // CHECK-USE-I386: @.offload_sizes.1 = private unnamed_addr constant [1 x i64] [i64 4]
96 // CHECK-USE-I386: @.offload_maptypes.2 = private unnamed_addr constant [1 x i64] [i64 [[#0x2427]]]
97 // CHECK-USE-I386: @.offload_sizes.3 = private unnamed_addr constant [3 x i64] [i64 0, i64 4, i64 4]
98 // CHECK-USE-I386: @.offload_maptypes.4 = private unnamed_addr constant [3 x i64] [i64 [[#0x2020]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]]]
99 //.
100 // CHECK-NOUSE-PPC64LE: @.offload_sizes = private unnamed_addr constant [7 x i64] [i64 0, i64 4, i64 4, i64 4, i64 0, i64 4, i64 4]
101 // CHECK-NOUSE-PPC64LE: @.offload_maptypes = private unnamed_addr constant [7 x i64] [i64 [[#0x2000]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]], i64 [[#0x2003]], i64 [[#0x2000]], i64 [[#0x5000000002003]], i64 [[#0x5000000002003]]]
102 // CHECK-NOUSE-PPC64LE: @.offload_sizes.1 = private unnamed_addr constant [1 x i64] [i64 4]
103 // CHECK-NOUSE-PPC64LE: @.offload_maptypes.2 = private unnamed_addr constant [1 x i64] [i64 [[#0x2407]]]
104 // CHECK-NOUSE-PPC64LE: @.offload_sizes.3 = private unnamed_addr constant [3 x i64] [i64 0, i64 4, i64 4]
105 // CHECK-NOUSE-PPC64LE: @.offload_maptypes.4 = private unnamed_addr constant [3 x i64] [i64 [[#0x2000]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]]]
106 //.
107 // CHECK-NOUSE-I386: @.offload_sizes = private unnamed_addr constant [7 x i64] [i64 0, i64 4, i64 4, i64 4, i64 0, i64 4, i64 4]
108 // CHECK-NOUSE-I386: @.offload_maptypes = private unnamed_addr constant [7 x i64] [i64 [[#0x2000]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]], i64 [[#0x2003]], i64 [[#0x2000]], i64 [[#0x5000000002003]], i64 [[#0x5000000002003]]]
109 // CHECK-NOUSE-I386: @.offload_sizes.1 = private unnamed_addr constant [1 x i64] [i64 4]
110 // CHECK-NOUSE-I386: @.offload_maptypes.2 = private unnamed_addr constant [1 x i64] [i64 [[#0x2407]]]
111 // CHECK-NOUSE-I386: @.offload_sizes.3 = private unnamed_addr constant [3 x i64] [i64 0, i64 4, i64 4]
112 // CHECK-NOUSE-I386: @.offload_maptypes.4 = private unnamed_addr constant [3 x i64] [i64 [[#0x2000]], i64 [[#0x1000000002003]], i64 [[#0x1000000002003]]]
113 //.
114 struct ST {
115 int i;
116 int j;
117 void test_present_members();
118 };
119
120 // CHECK-USE-PPC64LE-LABEL: @_Z20explicit_maps_singlei(
121 // CHECK-USE-PPC64LE-NEXT: entry:
122 // CHECK-USE-PPC64LE-NEXT: [[II_ADDR:%.*]] = alloca i32, align 4
123 // CHECK-USE-PPC64LE-NEXT: [[A:%.*]] = alloca i32, align 4
124 // CHECK-USE-PPC64LE-NEXT: [[ST1:%.*]] = alloca [[STRUCT_ST:%.*]], align 4
125 // CHECK-USE-PPC64LE-NEXT: [[ST2:%.*]] = alloca [[STRUCT_ST]], align 4
126 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [7 x i8*], align 8
127 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [7 x i8*], align 8
128 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [7 x i8*], align 8
129 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [7 x i64], align 8
130 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_BASEPTRS3:%.*]] = alloca [1 x i8*], align 8
131 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_PTRS4:%.*]] = alloca [1 x i8*], align 8
132 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_MAPPERS5:%.*]] = alloca [1 x i8*], align 8
133 // CHECK-USE-PPC64LE-NEXT: store i32 [[II:%.*]], i32* [[II_ADDR]], align 4
134 // CHECK-USE-PPC64LE-NEXT: [[TMP0:%.*]] = load i32, i32* [[II_ADDR]], align 4
135 // CHECK-USE-PPC64LE-NEXT: store i32 [[TMP0]], i32* [[A]], align 4
136 // CHECK-USE-PPC64LE-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 0
137 // CHECK-USE-PPC64LE-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 1
138 // CHECK-USE-PPC64LE-NEXT: [[TMP1:%.*]] = getelementptr i32, i32* [[J]], i32 1
139 // CHECK-USE-PPC64LE-NEXT: [[TMP2:%.*]] = bitcast i32* [[I]] to i8*
140 // CHECK-USE-PPC64LE-NEXT: [[TMP3:%.*]] = bitcast i32* [[TMP1]] to i8*
141 // CHECK-USE-PPC64LE-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP3]] to i64
142 // CHECK-USE-PPC64LE-NEXT: [[TMP5:%.*]] = ptrtoint i8* [[TMP2]] to i64
143 // CHECK-USE-PPC64LE-NEXT: [[TMP6:%.*]] = sub i64 [[TMP4]], [[TMP5]]
144 // CHECK-USE-PPC64LE-NEXT: [[TMP7:%.*]] = sdiv exact i64 [[TMP6]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
145 // CHECK-USE-PPC64LE-NEXT: [[I1:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 0
146 // CHECK-USE-PPC64LE-NEXT: [[J2:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 1
147 // CHECK-USE-PPC64LE-NEXT: [[TMP8:%.*]] = getelementptr i32, i32* [[J2]], i32 1
148 // CHECK-USE-PPC64LE-NEXT: [[TMP9:%.*]] = bitcast i32* [[I1]] to i8*
149 // CHECK-USE-PPC64LE-NEXT: [[TMP10:%.*]] = bitcast i32* [[TMP8]] to i8*
150 // CHECK-USE-PPC64LE-NEXT: [[TMP11:%.*]] = ptrtoint i8* [[TMP10]] to i64
151 // CHECK-USE-PPC64LE-NEXT: [[TMP12:%.*]] = ptrtoint i8* [[TMP9]] to i64
152 // CHECK-USE-PPC64LE-NEXT: [[TMP13:%.*]] = sub i64 [[TMP11]], [[TMP12]]
153 // CHECK-USE-PPC64LE-NEXT: [[TMP14:%.*]] = sdiv exact i64 [[TMP13]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
154 // CHECK-USE-PPC64LE-NEXT: [[TMP15:%.*]] = bitcast [7 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
155 // CHECK-USE-PPC64LE-NEXT: call void @llvm.memcpy.p0i8.p0i8.i64(i8* align 8 [[TMP15]], i8* align 8 bitcast ([7 x i64]* @.offload_sizes to i8*), i64 56, i1 false)
156 // CHECK-USE-PPC64LE-NEXT: [[TMP16:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
157 // CHECK-USE-PPC64LE-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to %struct.ST**
158 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP17]], align 8
159 // CHECK-USE-PPC64LE-NEXT: [[TMP18:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
160 // CHECK-USE-PPC64LE-NEXT: [[TMP19:%.*]] = bitcast i8** [[TMP18]] to i32**
161 // CHECK-USE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP19]], align 8
162 // CHECK-USE-PPC64LE-NEXT: [[TMP20:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
163 // CHECK-USE-PPC64LE-NEXT: store i64 [[TMP7]], i64* [[TMP20]], align 8
164 // CHECK-USE-PPC64LE-NEXT: [[TMP21:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0
165 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP21]], align 8
166 // CHECK-USE-PPC64LE-NEXT: [[TMP22:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
167 // CHECK-USE-PPC64LE-NEXT: [[TMP23:%.*]] = bitcast i8** [[TMP22]] to %struct.ST**
168 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP23]], align 8
169 // CHECK-USE-PPC64LE-NEXT: [[TMP24:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
170 // CHECK-USE-PPC64LE-NEXT: [[TMP25:%.*]] = bitcast i8** [[TMP24]] to i32**
171 // CHECK-USE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP25]], align 8
172 // CHECK-USE-PPC64LE-NEXT: [[TMP26:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1
173 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP26]], align 8
174 // CHECK-USE-PPC64LE-NEXT: [[TMP27:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
175 // CHECK-USE-PPC64LE-NEXT: [[TMP28:%.*]] = bitcast i8** [[TMP27]] to %struct.ST**
176 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP28]], align 8
177 // CHECK-USE-PPC64LE-NEXT: [[TMP29:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
178 // CHECK-USE-PPC64LE-NEXT: [[TMP30:%.*]] = bitcast i8** [[TMP29]] to i32**
179 // CHECK-USE-PPC64LE-NEXT: store i32* [[J]], i32** [[TMP30]], align 8
180 // CHECK-USE-PPC64LE-NEXT: [[TMP31:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2
181 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP31]], align 8
182 // CHECK-USE-PPC64LE-NEXT: [[TMP32:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3
183 // CHECK-USE-PPC64LE-NEXT: [[TMP33:%.*]] = bitcast i8** [[TMP32]] to i32**
184 // CHECK-USE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP33]], align 8
185 // CHECK-USE-PPC64LE-NEXT: [[TMP34:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 3
186 // CHECK-USE-PPC64LE-NEXT: [[TMP35:%.*]] = bitcast i8** [[TMP34]] to i32**
187 // CHECK-USE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP35]], align 8
188 // CHECK-USE-PPC64LE-NEXT: [[TMP36:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 3
189 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP36]], align 8
190 // CHECK-USE-PPC64LE-NEXT: [[TMP37:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 4
191 // CHECK-USE-PPC64LE-NEXT: [[TMP38:%.*]] = bitcast i8** [[TMP37]] to %struct.ST**
192 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP38]], align 8
193 // CHECK-USE-PPC64LE-NEXT: [[TMP39:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 4
194 // CHECK-USE-PPC64LE-NEXT: [[TMP40:%.*]] = bitcast i8** [[TMP39]] to i32**
195 // CHECK-USE-PPC64LE-NEXT: store i32* [[I1]], i32** [[TMP40]], align 8
196 // CHECK-USE-PPC64LE-NEXT: [[TMP41:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 4
197 // CHECK-USE-PPC64LE-NEXT: store i64 [[TMP14]], i64* [[TMP41]], align 8
198 // CHECK-USE-PPC64LE-NEXT: [[TMP42:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 4
199 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP42]], align 8
200 // CHECK-USE-PPC64LE-NEXT: [[TMP43:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 5
201 // CHECK-USE-PPC64LE-NEXT: [[TMP44:%.*]] = bitcast i8** [[TMP43]] to %struct.ST**
202 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP44]], align 8
203 // CHECK-USE-PPC64LE-NEXT: [[TMP45:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 5
204 // CHECK-USE-PPC64LE-NEXT: [[TMP46:%.*]] = bitcast i8** [[TMP45]] to i32**
205 // CHECK-USE-PPC64LE-NEXT: store i32* [[I1]], i32** [[TMP46]], align 8
206 // CHECK-USE-PPC64LE-NEXT: [[TMP47:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 5
207 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP47]], align 8
208 // CHECK-USE-PPC64LE-NEXT: [[TMP48:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 6
209 // CHECK-USE-PPC64LE-NEXT: [[TMP49:%.*]] = bitcast i8** [[TMP48]] to %struct.ST**
210 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP49]], align 8
211 // CHECK-USE-PPC64LE-NEXT: [[TMP50:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 6
212 // CHECK-USE-PPC64LE-NEXT: [[TMP51:%.*]] = bitcast i8** [[TMP50]] to i32**
213 // CHECK-USE-PPC64LE-NEXT: store i32* [[J2]], i32** [[TMP51]], align 8
214 // CHECK-USE-PPC64LE-NEXT: [[TMP52:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 6
215 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP52]], align 8
216 // CHECK-USE-PPC64LE-NEXT: [[TMP53:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
217 // CHECK-USE-PPC64LE-NEXT: [[TMP54:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
218 // CHECK-USE-PPC64LE-NEXT: [[TMP55:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
219 // CHECK-USE-PPC64LE-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
220 // CHECK-USE-PPC64LE-NEXT: [[TMP56:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
221 // CHECK-USE-PPC64LE-NEXT: store i32 1, i32* [[TMP56]], align 4
222 // CHECK-USE-PPC64LE-NEXT: [[TMP57:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
223 // CHECK-USE-PPC64LE-NEXT: store i32 7, i32* [[TMP57]], align 4
224 // CHECK-USE-PPC64LE-NEXT: [[TMP58:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
225 // CHECK-USE-PPC64LE-NEXT: store i8** [[TMP53]], i8*** [[TMP58]], align 8
226 // CHECK-USE-PPC64LE-NEXT: [[TMP59:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
227 // CHECK-USE-PPC64LE-NEXT: store i8** [[TMP54]], i8*** [[TMP59]], align 8
228 // CHECK-USE-PPC64LE-NEXT: [[TMP60:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
229 // CHECK-USE-PPC64LE-NEXT: store i64* [[TMP55]], i64** [[TMP60]], align 8
230 // CHECK-USE-PPC64LE-NEXT: [[TMP61:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
231 // CHECK-USE-PPC64LE-NEXT: store i64* getelementptr inbounds ([7 x i64], [7 x i64]* @.offload_maptypes, i32 0, i32 0), i64** [[TMP61]], align 8
232 // CHECK-USE-PPC64LE-NEXT: [[TMP62:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
233 // CHECK-USE-PPC64LE-NEXT: store i8** null, i8*** [[TMP62]], align 8
234 // CHECK-USE-PPC64LE-NEXT: [[TMP63:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
235 // CHECK-USE-PPC64LE-NEXT: store i8** null, i8*** [[TMP63]], align 8
236 // CHECK-USE-PPC64LE-NEXT: [[TMP64:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
237 // CHECK-USE-PPC64LE-NEXT: store i64 0, i64* [[TMP64]], align 8
238 // CHECK-USE-PPC64LE-NEXT: [[TMP65:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1:[0-9]+]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
239 // CHECK-USE-PPC64LE-NEXT: [[TMP66:%.*]] = icmp ne i32 [[TMP65]], 0
240 // CHECK-USE-PPC64LE-NEXT: br i1 [[TMP66]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
241 // CHECK-USE-PPC64LE: omp_offload.failed:
242 // CHECK-USE-PPC64LE-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782(%struct.ST* [[ST1]], i32* [[A]], %struct.ST* [[ST2]]) #[[ATTR3:[0-9]+]]
243 // CHECK-USE-PPC64LE-NEXT: br label [[OMP_OFFLOAD_CONT]]
244 // CHECK-USE-PPC64LE: omp_offload.cont:
245 // CHECK-USE-PPC64LE-NEXT: [[TMP67:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
246 // CHECK-USE-PPC64LE-NEXT: [[TMP68:%.*]] = bitcast i8** [[TMP67]] to i32**
247 // CHECK-USE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP68]], align 8
248 // CHECK-USE-PPC64LE-NEXT: [[TMP69:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
249 // CHECK-USE-PPC64LE-NEXT: [[TMP70:%.*]] = bitcast i8** [[TMP69]] to i32**
250 // CHECK-USE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP70]], align 8
251 // CHECK-USE-PPC64LE-NEXT: [[TMP71:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_MAPPERS5]], i64 0, i64 0
252 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP71]], align 8
253 // CHECK-USE-PPC64LE-NEXT: [[TMP72:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
254 // CHECK-USE-PPC64LE-NEXT: [[TMP73:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
255 // CHECK-USE-PPC64LE-NEXT: [[KERNEL_ARGS6:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8
256 // CHECK-USE-PPC64LE-NEXT: [[TMP74:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 0
257 // CHECK-USE-PPC64LE-NEXT: store i32 1, i32* [[TMP74]], align 4
258 // CHECK-USE-PPC64LE-NEXT: [[TMP75:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 1
259 // CHECK-USE-PPC64LE-NEXT: store i32 1, i32* [[TMP75]], align 4
260 // CHECK-USE-PPC64LE-NEXT: [[TMP76:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 2
261 // CHECK-USE-PPC64LE-NEXT: store i8** [[TMP72]], i8*** [[TMP76]], align 8
262 // CHECK-USE-PPC64LE-NEXT: [[TMP77:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 3
263 // CHECK-USE-PPC64LE-NEXT: store i8** [[TMP73]], i8*** [[TMP77]], align 8
264 // CHECK-USE-PPC64LE-NEXT: [[TMP78:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 4
265 // CHECK-USE-PPC64LE-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_sizes.1, i32 0, i32 0), i64** [[TMP78]], align 8
266 // CHECK-USE-PPC64LE-NEXT: [[TMP79:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 5
267 // CHECK-USE-PPC64LE-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_maptypes.2, i32 0, i32 0), i64** [[TMP79]], align 8
268 // CHECK-USE-PPC64LE-NEXT: [[TMP80:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 6
269 // CHECK-USE-PPC64LE-NEXT: store i8** null, i8*** [[TMP80]], align 8
270 // CHECK-USE-PPC64LE-NEXT: [[TMP81:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 7
271 // CHECK-USE-PPC64LE-NEXT: store i8** null, i8*** [[TMP81]], align 8
272 // CHECK-USE-PPC64LE-NEXT: [[TMP82:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 8
273 // CHECK-USE-PPC64LE-NEXT: store i64 0, i64* [[TMP82]], align 8
274 // CHECK-USE-PPC64LE-NEXT: [[TMP83:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]])
275 // CHECK-USE-PPC64LE-NEXT: [[TMP84:%.*]] = icmp ne i32 [[TMP83]], 0
276 // CHECK-USE-PPC64LE-NEXT: br i1 [[TMP84]], label [[OMP_OFFLOAD_FAILED7:%.*]], label [[OMP_OFFLOAD_CONT8:%.*]]
277 // CHECK-USE-PPC64LE: omp_offload.failed7:
278 // CHECK-USE-PPC64LE-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796(i32* [[A]]) #[[ATTR3]]
279 // CHECK-USE-PPC64LE-NEXT: br label [[OMP_OFFLOAD_CONT8]]
280 // CHECK-USE-PPC64LE: omp_offload.cont8:
281 // CHECK-USE-PPC64LE-NEXT: ret void
282 //
283 // CHECK-USE-I386-LABEL: @_Z20explicit_maps_singlei(
284 // CHECK-USE-I386-NEXT: entry:
285 // CHECK-USE-I386-NEXT: [[II_ADDR:%.*]] = alloca i32, align 4
286 // CHECK-USE-I386-NEXT: [[A:%.*]] = alloca i32, align 4
287 // CHECK-USE-I386-NEXT: [[ST1:%.*]] = alloca [[STRUCT_ST:%.*]], align 4
288 // CHECK-USE-I386-NEXT: [[ST2:%.*]] = alloca [[STRUCT_ST]], align 4
289 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [7 x i8*], align 4
290 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [7 x i8*], align 4
291 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [7 x i8*], align 4
292 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [7 x i64], align 4
293 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_BASEPTRS3:%.*]] = alloca [1 x i8*], align 4
294 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_PTRS4:%.*]] = alloca [1 x i8*], align 4
295 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_MAPPERS5:%.*]] = alloca [1 x i8*], align 4
296 // CHECK-USE-I386-NEXT: store i32 [[II:%.*]], i32* [[II_ADDR]], align 4
297 // CHECK-USE-I386-NEXT: [[TMP0:%.*]] = load i32, i32* [[II_ADDR]], align 4
298 // CHECK-USE-I386-NEXT: store i32 [[TMP0]], i32* [[A]], align 4
299 // CHECK-USE-I386-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 0
300 // CHECK-USE-I386-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 1
301 // CHECK-USE-I386-NEXT: [[TMP1:%.*]] = getelementptr i32, i32* [[J]], i32 1
302 // CHECK-USE-I386-NEXT: [[TMP2:%.*]] = bitcast i32* [[I]] to i8*
303 // CHECK-USE-I386-NEXT: [[TMP3:%.*]] = bitcast i32* [[TMP1]] to i8*
304 // CHECK-USE-I386-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP3]] to i64
305 // CHECK-USE-I386-NEXT: [[TMP5:%.*]] = ptrtoint i8* [[TMP2]] to i64
306 // CHECK-USE-I386-NEXT: [[TMP6:%.*]] = sub i64 [[TMP4]], [[TMP5]]
307 // CHECK-USE-I386-NEXT: [[TMP7:%.*]] = sdiv exact i64 [[TMP6]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
308 // CHECK-USE-I386-NEXT: [[I1:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 0
309 // CHECK-USE-I386-NEXT: [[J2:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 1
310 // CHECK-USE-I386-NEXT: [[TMP8:%.*]] = getelementptr i32, i32* [[J2]], i32 1
311 // CHECK-USE-I386-NEXT: [[TMP9:%.*]] = bitcast i32* [[I1]] to i8*
312 // CHECK-USE-I386-NEXT: [[TMP10:%.*]] = bitcast i32* [[TMP8]] to i8*
313 // CHECK-USE-I386-NEXT: [[TMP11:%.*]] = ptrtoint i8* [[TMP10]] to i64
314 // CHECK-USE-I386-NEXT: [[TMP12:%.*]] = ptrtoint i8* [[TMP9]] to i64
315 // CHECK-USE-I386-NEXT: [[TMP13:%.*]] = sub i64 [[TMP11]], [[TMP12]]
316 // CHECK-USE-I386-NEXT: [[TMP14:%.*]] = sdiv exact i64 [[TMP13]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
317 // CHECK-USE-I386-NEXT: [[TMP15:%.*]] = bitcast [7 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
318 // CHECK-USE-I386-NEXT: call void @llvm.memcpy.p0i8.p0i8.i32(i8* align 4 [[TMP15]], i8* align 4 bitcast ([7 x i64]* @.offload_sizes to i8*), i32 56, i1 false)
319 // CHECK-USE-I386-NEXT: [[TMP16:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
320 // CHECK-USE-I386-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to %struct.ST**
321 // CHECK-USE-I386-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP17]], align 4
322 // CHECK-USE-I386-NEXT: [[TMP18:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
323 // CHECK-USE-I386-NEXT: [[TMP19:%.*]] = bitcast i8** [[TMP18]] to i32**
324 // CHECK-USE-I386-NEXT: store i32* [[I]], i32** [[TMP19]], align 4
325 // CHECK-USE-I386-NEXT: [[TMP20:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
326 // CHECK-USE-I386-NEXT: store i64 [[TMP7]], i64* [[TMP20]], align 4
327 // CHECK-USE-I386-NEXT: [[TMP21:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0
328 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP21]], align 4
329 // CHECK-USE-I386-NEXT: [[TMP22:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
330 // CHECK-USE-I386-NEXT: [[TMP23:%.*]] = bitcast i8** [[TMP22]] to %struct.ST**
331 // CHECK-USE-I386-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP23]], align 4
332 // CHECK-USE-I386-NEXT: [[TMP24:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
333 // CHECK-USE-I386-NEXT: [[TMP25:%.*]] = bitcast i8** [[TMP24]] to i32**
334 // CHECK-USE-I386-NEXT: store i32* [[I]], i32** [[TMP25]], align 4
335 // CHECK-USE-I386-NEXT: [[TMP26:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 1
336 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP26]], align 4
337 // CHECK-USE-I386-NEXT: [[TMP27:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
338 // CHECK-USE-I386-NEXT: [[TMP28:%.*]] = bitcast i8** [[TMP27]] to %struct.ST**
339 // CHECK-USE-I386-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP28]], align 4
340 // CHECK-USE-I386-NEXT: [[TMP29:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
341 // CHECK-USE-I386-NEXT: [[TMP30:%.*]] = bitcast i8** [[TMP29]] to i32**
342 // CHECK-USE-I386-NEXT: store i32* [[J]], i32** [[TMP30]], align 4
343 // CHECK-USE-I386-NEXT: [[TMP31:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 2
344 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP31]], align 4
345 // CHECK-USE-I386-NEXT: [[TMP32:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3
346 // CHECK-USE-I386-NEXT: [[TMP33:%.*]] = bitcast i8** [[TMP32]] to i32**
347 // CHECK-USE-I386-NEXT: store i32* [[A]], i32** [[TMP33]], align 4
348 // CHECK-USE-I386-NEXT: [[TMP34:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 3
349 // CHECK-USE-I386-NEXT: [[TMP35:%.*]] = bitcast i8** [[TMP34]] to i32**
350 // CHECK-USE-I386-NEXT: store i32* [[A]], i32** [[TMP35]], align 4
351 // CHECK-USE-I386-NEXT: [[TMP36:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 3
352 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP36]], align 4
353 // CHECK-USE-I386-NEXT: [[TMP37:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 4
354 // CHECK-USE-I386-NEXT: [[TMP38:%.*]] = bitcast i8** [[TMP37]] to %struct.ST**
355 // CHECK-USE-I386-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP38]], align 4
356 // CHECK-USE-I386-NEXT: [[TMP39:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 4
357 // CHECK-USE-I386-NEXT: [[TMP40:%.*]] = bitcast i8** [[TMP39]] to i32**
358 // CHECK-USE-I386-NEXT: store i32* [[I1]], i32** [[TMP40]], align 4
359 // CHECK-USE-I386-NEXT: [[TMP41:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 4
360 // CHECK-USE-I386-NEXT: store i64 [[TMP14]], i64* [[TMP41]], align 4
361 // CHECK-USE-I386-NEXT: [[TMP42:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 4
362 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP42]], align 4
363 // CHECK-USE-I386-NEXT: [[TMP43:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 5
364 // CHECK-USE-I386-NEXT: [[TMP44:%.*]] = bitcast i8** [[TMP43]] to %struct.ST**
365 // CHECK-USE-I386-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP44]], align 4
366 // CHECK-USE-I386-NEXT: [[TMP45:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 5
367 // CHECK-USE-I386-NEXT: [[TMP46:%.*]] = bitcast i8** [[TMP45]] to i32**
368 // CHECK-USE-I386-NEXT: store i32* [[I1]], i32** [[TMP46]], align 4
369 // CHECK-USE-I386-NEXT: [[TMP47:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 5
370 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP47]], align 4
371 // CHECK-USE-I386-NEXT: [[TMP48:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 6
372 // CHECK-USE-I386-NEXT: [[TMP49:%.*]] = bitcast i8** [[TMP48]] to %struct.ST**
373 // CHECK-USE-I386-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP49]], align 4
374 // CHECK-USE-I386-NEXT: [[TMP50:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 6
375 // CHECK-USE-I386-NEXT: [[TMP51:%.*]] = bitcast i8** [[TMP50]] to i32**
376 // CHECK-USE-I386-NEXT: store i32* [[J2]], i32** [[TMP51]], align 4
377 // CHECK-USE-I386-NEXT: [[TMP52:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 6
378 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP52]], align 4
379 // CHECK-USE-I386-NEXT: [[TMP53:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
380 // CHECK-USE-I386-NEXT: [[TMP54:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
381 // CHECK-USE-I386-NEXT: [[TMP55:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
382 // CHECK-USE-I386-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
383 // CHECK-USE-I386-NEXT: [[TMP56:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
384 // CHECK-USE-I386-NEXT: store i32 1, i32* [[TMP56]], align 4
385 // CHECK-USE-I386-NEXT: [[TMP57:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
386 // CHECK-USE-I386-NEXT: store i32 7, i32* [[TMP57]], align 4
387 // CHECK-USE-I386-NEXT: [[TMP58:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
388 // CHECK-USE-I386-NEXT: store i8** [[TMP53]], i8*** [[TMP58]], align 4
389 // CHECK-USE-I386-NEXT: [[TMP59:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
390 // CHECK-USE-I386-NEXT: store i8** [[TMP54]], i8*** [[TMP59]], align 4
391 // CHECK-USE-I386-NEXT: [[TMP60:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
392 // CHECK-USE-I386-NEXT: store i64* [[TMP55]], i64** [[TMP60]], align 4
393 // CHECK-USE-I386-NEXT: [[TMP61:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
394 // CHECK-USE-I386-NEXT: store i64* getelementptr inbounds ([7 x i64], [7 x i64]* @.offload_maptypes, i32 0, i32 0), i64** [[TMP61]], align 4
395 // CHECK-USE-I386-NEXT: [[TMP62:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
396 // CHECK-USE-I386-NEXT: store i8** null, i8*** [[TMP62]], align 4
397 // CHECK-USE-I386-NEXT: [[TMP63:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
398 // CHECK-USE-I386-NEXT: store i8** null, i8*** [[TMP63]], align 4
399 // CHECK-USE-I386-NEXT: [[TMP64:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
400 // CHECK-USE-I386-NEXT: store i64 0, i64* [[TMP64]], align 8
401 // CHECK-USE-I386-NEXT: [[TMP65:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1:[0-9]+]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
402 // CHECK-USE-I386-NEXT: [[TMP66:%.*]] = icmp ne i32 [[TMP65]], 0
403 // CHECK-USE-I386-NEXT: br i1 [[TMP66]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
404 // CHECK-USE-I386: omp_offload.failed:
405 // CHECK-USE-I386-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782(%struct.ST* [[ST1]], i32* [[A]], %struct.ST* [[ST2]]) #[[ATTR3:[0-9]+]]
406 // CHECK-USE-I386-NEXT: br label [[OMP_OFFLOAD_CONT]]
407 // CHECK-USE-I386: omp_offload.cont:
408 // CHECK-USE-I386-NEXT: [[TMP67:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
409 // CHECK-USE-I386-NEXT: [[TMP68:%.*]] = bitcast i8** [[TMP67]] to i32**
410 // CHECK-USE-I386-NEXT: store i32* [[A]], i32** [[TMP68]], align 4
411 // CHECK-USE-I386-NEXT: [[TMP69:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
412 // CHECK-USE-I386-NEXT: [[TMP70:%.*]] = bitcast i8** [[TMP69]] to i32**
413 // CHECK-USE-I386-NEXT: store i32* [[A]], i32** [[TMP70]], align 4
414 // CHECK-USE-I386-NEXT: [[TMP71:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_MAPPERS5]], i32 0, i32 0
415 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP71]], align 4
416 // CHECK-USE-I386-NEXT: [[TMP72:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
417 // CHECK-USE-I386-NEXT: [[TMP73:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
418 // CHECK-USE-I386-NEXT: [[KERNEL_ARGS6:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8
419 // CHECK-USE-I386-NEXT: [[TMP74:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 0
420 // CHECK-USE-I386-NEXT: store i32 1, i32* [[TMP74]], align 4
421 // CHECK-USE-I386-NEXT: [[TMP75:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 1
422 // CHECK-USE-I386-NEXT: store i32 1, i32* [[TMP75]], align 4
423 // CHECK-USE-I386-NEXT: [[TMP76:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 2
424 // CHECK-USE-I386-NEXT: store i8** [[TMP72]], i8*** [[TMP76]], align 4
425 // CHECK-USE-I386-NEXT: [[TMP77:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 3
426 // CHECK-USE-I386-NEXT: store i8** [[TMP73]], i8*** [[TMP77]], align 4
427 // CHECK-USE-I386-NEXT: [[TMP78:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 4
428 // CHECK-USE-I386-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_sizes.1, i32 0, i32 0), i64** [[TMP78]], align 4
429 // CHECK-USE-I386-NEXT: [[TMP79:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 5
430 // CHECK-USE-I386-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_maptypes.2, i32 0, i32 0), i64** [[TMP79]], align 4
431 // CHECK-USE-I386-NEXT: [[TMP80:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 6
432 // CHECK-USE-I386-NEXT: store i8** null, i8*** [[TMP80]], align 4
433 // CHECK-USE-I386-NEXT: [[TMP81:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 7
434 // CHECK-USE-I386-NEXT: store i8** null, i8*** [[TMP81]], align 4
435 // CHECK-USE-I386-NEXT: [[TMP82:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 8
436 // CHECK-USE-I386-NEXT: store i64 0, i64* [[TMP82]], align 8
437 // CHECK-USE-I386-NEXT: [[TMP83:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]])
438 // CHECK-USE-I386-NEXT: [[TMP84:%.*]] = icmp ne i32 [[TMP83]], 0
439 // CHECK-USE-I386-NEXT: br i1 [[TMP84]], label [[OMP_OFFLOAD_FAILED7:%.*]], label [[OMP_OFFLOAD_CONT8:%.*]]
440 // CHECK-USE-I386: omp_offload.failed7:
441 // CHECK-USE-I386-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796(i32* [[A]]) #[[ATTR3]]
442 // CHECK-USE-I386-NEXT: br label [[OMP_OFFLOAD_CONT8]]
443 // CHECK-USE-I386: omp_offload.cont8:
444 // CHECK-USE-I386-NEXT: ret void
445 //
446 // CHECK-NOUSE-PPC64LE-LABEL: @_Z20explicit_maps_singlei(
447 // CHECK-NOUSE-PPC64LE-NEXT: entry:
448 // CHECK-NOUSE-PPC64LE-NEXT: [[II_ADDR:%.*]] = alloca i32, align 4
449 // CHECK-NOUSE-PPC64LE-NEXT: [[A:%.*]] = alloca i32, align 4
450 // CHECK-NOUSE-PPC64LE-NEXT: [[ST1:%.*]] = alloca [[STRUCT_ST:%.*]], align 4
451 // CHECK-NOUSE-PPC64LE-NEXT: [[ST2:%.*]] = alloca [[STRUCT_ST]], align 4
452 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [7 x i8*], align 8
453 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [7 x i8*], align 8
454 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [7 x i8*], align 8
455 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [7 x i64], align 8
456 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_BASEPTRS3:%.*]] = alloca [1 x i8*], align 8
457 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_PTRS4:%.*]] = alloca [1 x i8*], align 8
458 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_MAPPERS5:%.*]] = alloca [1 x i8*], align 8
459 // CHECK-NOUSE-PPC64LE-NEXT: store i32 [[II:%.*]], i32* [[II_ADDR]], align 4
460 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP0:%.*]] = load i32, i32* [[II_ADDR]], align 4
461 // CHECK-NOUSE-PPC64LE-NEXT: store i32 [[TMP0]], i32* [[A]], align 4
462 // CHECK-NOUSE-PPC64LE-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 0
463 // CHECK-NOUSE-PPC64LE-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 1
464 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP1:%.*]] = getelementptr i32, i32* [[J]], i32 1
465 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP2:%.*]] = bitcast i32* [[I]] to i8*
466 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP3:%.*]] = bitcast i32* [[TMP1]] to i8*
467 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP3]] to i64
468 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP5:%.*]] = ptrtoint i8* [[TMP2]] to i64
469 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP6:%.*]] = sub i64 [[TMP4]], [[TMP5]]
470 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP7:%.*]] = sdiv exact i64 [[TMP6]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
471 // CHECK-NOUSE-PPC64LE-NEXT: [[I1:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 0
472 // CHECK-NOUSE-PPC64LE-NEXT: [[J2:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 1
473 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP8:%.*]] = getelementptr i32, i32* [[J2]], i32 1
474 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP9:%.*]] = bitcast i32* [[I1]] to i8*
475 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP10:%.*]] = bitcast i32* [[TMP8]] to i8*
476 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP11:%.*]] = ptrtoint i8* [[TMP10]] to i64
477 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP12:%.*]] = ptrtoint i8* [[TMP9]] to i64
478 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP13:%.*]] = sub i64 [[TMP11]], [[TMP12]]
479 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP14:%.*]] = sdiv exact i64 [[TMP13]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
480 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP15:%.*]] = bitcast [7 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
481 // CHECK-NOUSE-PPC64LE-NEXT: call void @llvm.memcpy.p0i8.p0i8.i64(i8* align 8 [[TMP15]], i8* align 8 bitcast ([7 x i64]* @.offload_sizes to i8*), i64 56, i1 false)
482 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP16:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
483 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to %struct.ST**
484 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP17]], align 8
485 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP18:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
486 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP19:%.*]] = bitcast i8** [[TMP18]] to i32**
487 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP19]], align 8
488 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP20:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
489 // CHECK-NOUSE-PPC64LE-NEXT: store i64 [[TMP7]], i64* [[TMP20]], align 8
490 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP21:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0
491 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP21]], align 8
492 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP22:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
493 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP23:%.*]] = bitcast i8** [[TMP22]] to %struct.ST**
494 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP23]], align 8
495 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP24:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
496 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP25:%.*]] = bitcast i8** [[TMP24]] to i32**
497 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP25]], align 8
498 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP26:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1
499 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP26]], align 8
500 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP27:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
501 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP28:%.*]] = bitcast i8** [[TMP27]] to %struct.ST**
502 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP28]], align 8
503 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP29:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
504 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP30:%.*]] = bitcast i8** [[TMP29]] to i32**
505 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[J]], i32** [[TMP30]], align 8
506 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP31:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2
507 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP31]], align 8
508 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP32:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3
509 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP33:%.*]] = bitcast i8** [[TMP32]] to i32**
510 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP33]], align 8
511 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP34:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 3
512 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP35:%.*]] = bitcast i8** [[TMP34]] to i32**
513 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP35]], align 8
514 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP36:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 3
515 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP36]], align 8
516 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP37:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 4
517 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP38:%.*]] = bitcast i8** [[TMP37]] to %struct.ST**
518 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP38]], align 8
519 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP39:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 4
520 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP40:%.*]] = bitcast i8** [[TMP39]] to i32**
521 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[I1]], i32** [[TMP40]], align 8
522 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP41:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 4
523 // CHECK-NOUSE-PPC64LE-NEXT: store i64 [[TMP14]], i64* [[TMP41]], align 8
524 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP42:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 4
525 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP42]], align 8
526 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP43:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 5
527 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP44:%.*]] = bitcast i8** [[TMP43]] to %struct.ST**
528 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP44]], align 8
529 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP45:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 5
530 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP46:%.*]] = bitcast i8** [[TMP45]] to i32**
531 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[I1]], i32** [[TMP46]], align 8
532 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP47:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 5
533 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP47]], align 8
534 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP48:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 6
535 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP49:%.*]] = bitcast i8** [[TMP48]] to %struct.ST**
536 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP49]], align 8
537 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP50:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 6
538 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP51:%.*]] = bitcast i8** [[TMP50]] to i32**
539 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[J2]], i32** [[TMP51]], align 8
540 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP52:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 6
541 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP52]], align 8
542 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP53:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
543 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP54:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
544 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP55:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
545 // CHECK-NOUSE-PPC64LE-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
546 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP56:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
547 // CHECK-NOUSE-PPC64LE-NEXT: store i32 1, i32* [[TMP56]], align 4
548 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP57:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
549 // CHECK-NOUSE-PPC64LE-NEXT: store i32 7, i32* [[TMP57]], align 4
550 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP58:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
551 // CHECK-NOUSE-PPC64LE-NEXT: store i8** [[TMP53]], i8*** [[TMP58]], align 8
552 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP59:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
553 // CHECK-NOUSE-PPC64LE-NEXT: store i8** [[TMP54]], i8*** [[TMP59]], align 8
554 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP60:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
555 // CHECK-NOUSE-PPC64LE-NEXT: store i64* [[TMP55]], i64** [[TMP60]], align 8
556 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP61:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
557 // CHECK-NOUSE-PPC64LE-NEXT: store i64* getelementptr inbounds ([7 x i64], [7 x i64]* @.offload_maptypes, i32 0, i32 0), i64** [[TMP61]], align 8
558 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP62:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
559 // CHECK-NOUSE-PPC64LE-NEXT: store i8** null, i8*** [[TMP62]], align 8
560 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP63:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
561 // CHECK-NOUSE-PPC64LE-NEXT: store i8** null, i8*** [[TMP63]], align 8
562 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP64:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
563 // CHECK-NOUSE-PPC64LE-NEXT: store i64 0, i64* [[TMP64]], align 8
564 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP65:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1:[0-9]+]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
565 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP66:%.*]] = icmp ne i32 [[TMP65]], 0
566 // CHECK-NOUSE-PPC64LE-NEXT: br i1 [[TMP66]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
567 // CHECK-NOUSE-PPC64LE: omp_offload.failed:
568 // CHECK-NOUSE-PPC64LE-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782() #[[ATTR3:[0-9]+]]
569 // CHECK-NOUSE-PPC64LE-NEXT: br label [[OMP_OFFLOAD_CONT]]
570 // CHECK-NOUSE-PPC64LE: omp_offload.cont:
571 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP67:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
572 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP68:%.*]] = bitcast i8** [[TMP67]] to i32**
573 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP68]], align 8
574 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP69:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
575 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP70:%.*]] = bitcast i8** [[TMP69]] to i32**
576 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[A]], i32** [[TMP70]], align 8
577 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP71:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_MAPPERS5]], i64 0, i64 0
578 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP71]], align 8
579 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP72:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
580 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP73:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
581 // CHECK-NOUSE-PPC64LE-NEXT: [[KERNEL_ARGS6:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8
582 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP74:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 0
583 // CHECK-NOUSE-PPC64LE-NEXT: store i32 1, i32* [[TMP74]], align 4
584 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP75:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 1
585 // CHECK-NOUSE-PPC64LE-NEXT: store i32 1, i32* [[TMP75]], align 4
586 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP76:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 2
587 // CHECK-NOUSE-PPC64LE-NEXT: store i8** [[TMP72]], i8*** [[TMP76]], align 8
588 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP77:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 3
589 // CHECK-NOUSE-PPC64LE-NEXT: store i8** [[TMP73]], i8*** [[TMP77]], align 8
590 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP78:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 4
591 // CHECK-NOUSE-PPC64LE-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_sizes.1, i32 0, i32 0), i64** [[TMP78]], align 8
592 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP79:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 5
593 // CHECK-NOUSE-PPC64LE-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_maptypes.2, i32 0, i32 0), i64** [[TMP79]], align 8
594 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP80:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 6
595 // CHECK-NOUSE-PPC64LE-NEXT: store i8** null, i8*** [[TMP80]], align 8
596 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP81:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 7
597 // CHECK-NOUSE-PPC64LE-NEXT: store i8** null, i8*** [[TMP81]], align 8
598 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP82:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 8
599 // CHECK-NOUSE-PPC64LE-NEXT: store i64 0, i64* [[TMP82]], align 8
600 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP83:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]])
601 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP84:%.*]] = icmp ne i32 [[TMP83]], 0
602 // CHECK-NOUSE-PPC64LE-NEXT: br i1 [[TMP84]], label [[OMP_OFFLOAD_FAILED7:%.*]], label [[OMP_OFFLOAD_CONT8:%.*]]
603 // CHECK-NOUSE-PPC64LE: omp_offload.failed7:
604 // CHECK-NOUSE-PPC64LE-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796() #[[ATTR3]]
605 // CHECK-NOUSE-PPC64LE-NEXT: br label [[OMP_OFFLOAD_CONT8]]
606 // CHECK-NOUSE-PPC64LE: omp_offload.cont8:
607 // CHECK-NOUSE-PPC64LE-NEXT: ret void
608 //
609 // CHECK-NOUSE-I386-LABEL: @_Z20explicit_maps_singlei(
610 // CHECK-NOUSE-I386-NEXT: entry:
611 // CHECK-NOUSE-I386-NEXT: [[II_ADDR:%.*]] = alloca i32, align 4
612 // CHECK-NOUSE-I386-NEXT: [[A:%.*]] = alloca i32, align 4
613 // CHECK-NOUSE-I386-NEXT: [[ST1:%.*]] = alloca [[STRUCT_ST:%.*]], align 4
614 // CHECK-NOUSE-I386-NEXT: [[ST2:%.*]] = alloca [[STRUCT_ST]], align 4
615 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [7 x i8*], align 4
616 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [7 x i8*], align 4
617 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [7 x i8*], align 4
618 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [7 x i64], align 4
619 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_BASEPTRS3:%.*]] = alloca [1 x i8*], align 4
620 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_PTRS4:%.*]] = alloca [1 x i8*], align 4
621 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_MAPPERS5:%.*]] = alloca [1 x i8*], align 4
622 // CHECK-NOUSE-I386-NEXT: store i32 [[II:%.*]], i32* [[II_ADDR]], align 4
623 // CHECK-NOUSE-I386-NEXT: [[TMP0:%.*]] = load i32, i32* [[II_ADDR]], align 4
624 // CHECK-NOUSE-I386-NEXT: store i32 [[TMP0]], i32* [[A]], align 4
625 // CHECK-NOUSE-I386-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 0
626 // CHECK-NOUSE-I386-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST1]], i32 0, i32 1
627 // CHECK-NOUSE-I386-NEXT: [[TMP1:%.*]] = getelementptr i32, i32* [[J]], i32 1
628 // CHECK-NOUSE-I386-NEXT: [[TMP2:%.*]] = bitcast i32* [[I]] to i8*
629 // CHECK-NOUSE-I386-NEXT: [[TMP3:%.*]] = bitcast i32* [[TMP1]] to i8*
630 // CHECK-NOUSE-I386-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP3]] to i64
631 // CHECK-NOUSE-I386-NEXT: [[TMP5:%.*]] = ptrtoint i8* [[TMP2]] to i64
632 // CHECK-NOUSE-I386-NEXT: [[TMP6:%.*]] = sub i64 [[TMP4]], [[TMP5]]
633 // CHECK-NOUSE-I386-NEXT: [[TMP7:%.*]] = sdiv exact i64 [[TMP6]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
634 // CHECK-NOUSE-I386-NEXT: [[I1:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 0
635 // CHECK-NOUSE-I386-NEXT: [[J2:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[ST2]], i32 0, i32 1
636 // CHECK-NOUSE-I386-NEXT: [[TMP8:%.*]] = getelementptr i32, i32* [[J2]], i32 1
637 // CHECK-NOUSE-I386-NEXT: [[TMP9:%.*]] = bitcast i32* [[I1]] to i8*
638 // CHECK-NOUSE-I386-NEXT: [[TMP10:%.*]] = bitcast i32* [[TMP8]] to i8*
639 // CHECK-NOUSE-I386-NEXT: [[TMP11:%.*]] = ptrtoint i8* [[TMP10]] to i64
640 // CHECK-NOUSE-I386-NEXT: [[TMP12:%.*]] = ptrtoint i8* [[TMP9]] to i64
641 // CHECK-NOUSE-I386-NEXT: [[TMP13:%.*]] = sub i64 [[TMP11]], [[TMP12]]
642 // CHECK-NOUSE-I386-NEXT: [[TMP14:%.*]] = sdiv exact i64 [[TMP13]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
643 // CHECK-NOUSE-I386-NEXT: [[TMP15:%.*]] = bitcast [7 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
644 // CHECK-NOUSE-I386-NEXT: call void @llvm.memcpy.p0i8.p0i8.i32(i8* align 4 [[TMP15]], i8* align 4 bitcast ([7 x i64]* @.offload_sizes to i8*), i32 56, i1 false)
645 // CHECK-NOUSE-I386-NEXT: [[TMP16:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
646 // CHECK-NOUSE-I386-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to %struct.ST**
647 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP17]], align 4
648 // CHECK-NOUSE-I386-NEXT: [[TMP18:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
649 // CHECK-NOUSE-I386-NEXT: [[TMP19:%.*]] = bitcast i8** [[TMP18]] to i32**
650 // CHECK-NOUSE-I386-NEXT: store i32* [[I]], i32** [[TMP19]], align 4
651 // CHECK-NOUSE-I386-NEXT: [[TMP20:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
652 // CHECK-NOUSE-I386-NEXT: store i64 [[TMP7]], i64* [[TMP20]], align 4
653 // CHECK-NOUSE-I386-NEXT: [[TMP21:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0
654 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP21]], align 4
655 // CHECK-NOUSE-I386-NEXT: [[TMP22:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
656 // CHECK-NOUSE-I386-NEXT: [[TMP23:%.*]] = bitcast i8** [[TMP22]] to %struct.ST**
657 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP23]], align 4
658 // CHECK-NOUSE-I386-NEXT: [[TMP24:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
659 // CHECK-NOUSE-I386-NEXT: [[TMP25:%.*]] = bitcast i8** [[TMP24]] to i32**
660 // CHECK-NOUSE-I386-NEXT: store i32* [[I]], i32** [[TMP25]], align 4
661 // CHECK-NOUSE-I386-NEXT: [[TMP26:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 1
662 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP26]], align 4
663 // CHECK-NOUSE-I386-NEXT: [[TMP27:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
664 // CHECK-NOUSE-I386-NEXT: [[TMP28:%.*]] = bitcast i8** [[TMP27]] to %struct.ST**
665 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[ST1]], %struct.ST** [[TMP28]], align 4
666 // CHECK-NOUSE-I386-NEXT: [[TMP29:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
667 // CHECK-NOUSE-I386-NEXT: [[TMP30:%.*]] = bitcast i8** [[TMP29]] to i32**
668 // CHECK-NOUSE-I386-NEXT: store i32* [[J]], i32** [[TMP30]], align 4
669 // CHECK-NOUSE-I386-NEXT: [[TMP31:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 2
670 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP31]], align 4
671 // CHECK-NOUSE-I386-NEXT: [[TMP32:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 3
672 // CHECK-NOUSE-I386-NEXT: [[TMP33:%.*]] = bitcast i8** [[TMP32]] to i32**
673 // CHECK-NOUSE-I386-NEXT: store i32* [[A]], i32** [[TMP33]], align 4
674 // CHECK-NOUSE-I386-NEXT: [[TMP34:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 3
675 // CHECK-NOUSE-I386-NEXT: [[TMP35:%.*]] = bitcast i8** [[TMP34]] to i32**
676 // CHECK-NOUSE-I386-NEXT: store i32* [[A]], i32** [[TMP35]], align 4
677 // CHECK-NOUSE-I386-NEXT: [[TMP36:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 3
678 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP36]], align 4
679 // CHECK-NOUSE-I386-NEXT: [[TMP37:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 4
680 // CHECK-NOUSE-I386-NEXT: [[TMP38:%.*]] = bitcast i8** [[TMP37]] to %struct.ST**
681 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP38]], align 4
682 // CHECK-NOUSE-I386-NEXT: [[TMP39:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 4
683 // CHECK-NOUSE-I386-NEXT: [[TMP40:%.*]] = bitcast i8** [[TMP39]] to i32**
684 // CHECK-NOUSE-I386-NEXT: store i32* [[I1]], i32** [[TMP40]], align 4
685 // CHECK-NOUSE-I386-NEXT: [[TMP41:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 4
686 // CHECK-NOUSE-I386-NEXT: store i64 [[TMP14]], i64* [[TMP41]], align 4
687 // CHECK-NOUSE-I386-NEXT: [[TMP42:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 4
688 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP42]], align 4
689 // CHECK-NOUSE-I386-NEXT: [[TMP43:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 5
690 // CHECK-NOUSE-I386-NEXT: [[TMP44:%.*]] = bitcast i8** [[TMP43]] to %struct.ST**
691 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP44]], align 4
692 // CHECK-NOUSE-I386-NEXT: [[TMP45:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 5
693 // CHECK-NOUSE-I386-NEXT: [[TMP46:%.*]] = bitcast i8** [[TMP45]] to i32**
694 // CHECK-NOUSE-I386-NEXT: store i32* [[I1]], i32** [[TMP46]], align 4
695 // CHECK-NOUSE-I386-NEXT: [[TMP47:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 5
696 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP47]], align 4
697 // CHECK-NOUSE-I386-NEXT: [[TMP48:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 6
698 // CHECK-NOUSE-I386-NEXT: [[TMP49:%.*]] = bitcast i8** [[TMP48]] to %struct.ST**
699 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[ST2]], %struct.ST** [[TMP49]], align 4
700 // CHECK-NOUSE-I386-NEXT: [[TMP50:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 6
701 // CHECK-NOUSE-I386-NEXT: [[TMP51:%.*]] = bitcast i8** [[TMP50]] to i32**
702 // CHECK-NOUSE-I386-NEXT: store i32* [[J2]], i32** [[TMP51]], align 4
703 // CHECK-NOUSE-I386-NEXT: [[TMP52:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 6
704 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP52]], align 4
705 // CHECK-NOUSE-I386-NEXT: [[TMP53:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
706 // CHECK-NOUSE-I386-NEXT: [[TMP54:%.*]] = getelementptr inbounds [7 x i8*], [7 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
707 // CHECK-NOUSE-I386-NEXT: [[TMP55:%.*]] = getelementptr inbounds [7 x i64], [7 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
708 // CHECK-NOUSE-I386-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
709 // CHECK-NOUSE-I386-NEXT: [[TMP56:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
710 // CHECK-NOUSE-I386-NEXT: store i32 1, i32* [[TMP56]], align 4
711 // CHECK-NOUSE-I386-NEXT: [[TMP57:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
712 // CHECK-NOUSE-I386-NEXT: store i32 7, i32* [[TMP57]], align 4
713 // CHECK-NOUSE-I386-NEXT: [[TMP58:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
714 // CHECK-NOUSE-I386-NEXT: store i8** [[TMP53]], i8*** [[TMP58]], align 4
715 // CHECK-NOUSE-I386-NEXT: [[TMP59:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
716 // CHECK-NOUSE-I386-NEXT: store i8** [[TMP54]], i8*** [[TMP59]], align 4
717 // CHECK-NOUSE-I386-NEXT: [[TMP60:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
718 // CHECK-NOUSE-I386-NEXT: store i64* [[TMP55]], i64** [[TMP60]], align 4
719 // CHECK-NOUSE-I386-NEXT: [[TMP61:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
720 // CHECK-NOUSE-I386-NEXT: store i64* getelementptr inbounds ([7 x i64], [7 x i64]* @.offload_maptypes, i32 0, i32 0), i64** [[TMP61]], align 4
721 // CHECK-NOUSE-I386-NEXT: [[TMP62:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
722 // CHECK-NOUSE-I386-NEXT: store i8** null, i8*** [[TMP62]], align 4
723 // CHECK-NOUSE-I386-NEXT: [[TMP63:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
724 // CHECK-NOUSE-I386-NEXT: store i8** null, i8*** [[TMP63]], align 4
725 // CHECK-NOUSE-I386-NEXT: [[TMP64:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
726 // CHECK-NOUSE-I386-NEXT: store i64 0, i64* [[TMP64]], align 8
727 // CHECK-NOUSE-I386-NEXT: [[TMP65:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1:[0-9]+]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
728 // CHECK-NOUSE-I386-NEXT: [[TMP66:%.*]] = icmp ne i32 [[TMP65]], 0
729 // CHECK-NOUSE-I386-NEXT: br i1 [[TMP66]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
730 // CHECK-NOUSE-I386: omp_offload.failed:
731 // CHECK-NOUSE-I386-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l782() #[[ATTR3:[0-9]+]]
732 // CHECK-NOUSE-I386-NEXT: br label [[OMP_OFFLOAD_CONT]]
733 // CHECK-NOUSE-I386: omp_offload.cont:
734 // CHECK-NOUSE-I386-NEXT: [[TMP67:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
735 // CHECK-NOUSE-I386-NEXT: [[TMP68:%.*]] = bitcast i8** [[TMP67]] to i32**
736 // CHECK-NOUSE-I386-NEXT: store i32* [[A]], i32** [[TMP68]], align 4
737 // CHECK-NOUSE-I386-NEXT: [[TMP69:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
738 // CHECK-NOUSE-I386-NEXT: [[TMP70:%.*]] = bitcast i8** [[TMP69]] to i32**
739 // CHECK-NOUSE-I386-NEXT: store i32* [[A]], i32** [[TMP70]], align 4
740 // CHECK-NOUSE-I386-NEXT: [[TMP71:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_MAPPERS5]], i32 0, i32 0
741 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP71]], align 4
742 // CHECK-NOUSE-I386-NEXT: [[TMP72:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_BASEPTRS3]], i32 0, i32 0
743 // CHECK-NOUSE-I386-NEXT: [[TMP73:%.*]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[DOTOFFLOAD_PTRS4]], i32 0, i32 0
744 // CHECK-NOUSE-I386-NEXT: [[KERNEL_ARGS6:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS]], align 8
745 // CHECK-NOUSE-I386-NEXT: [[TMP74:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 0
746 // CHECK-NOUSE-I386-NEXT: store i32 1, i32* [[TMP74]], align 4
747 // CHECK-NOUSE-I386-NEXT: [[TMP75:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 1
748 // CHECK-NOUSE-I386-NEXT: store i32 1, i32* [[TMP75]], align 4
749 // CHECK-NOUSE-I386-NEXT: [[TMP76:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 2
750 // CHECK-NOUSE-I386-NEXT: store i8** [[TMP72]], i8*** [[TMP76]], align 4
751 // CHECK-NOUSE-I386-NEXT: [[TMP77:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 3
752 // CHECK-NOUSE-I386-NEXT: store i8** [[TMP73]], i8*** [[TMP77]], align 4
753 // CHECK-NOUSE-I386-NEXT: [[TMP78:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 4
754 // CHECK-NOUSE-I386-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_sizes.1, i32 0, i32 0), i64** [[TMP78]], align 4
755 // CHECK-NOUSE-I386-NEXT: [[TMP79:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 5
756 // CHECK-NOUSE-I386-NEXT: store i64* getelementptr inbounds ([1 x i64], [1 x i64]* @.offload_maptypes.2, i32 0, i32 0), i64** [[TMP79]], align 4
757 // CHECK-NOUSE-I386-NEXT: [[TMP80:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 6
758 // CHECK-NOUSE-I386-NEXT: store i8** null, i8*** [[TMP80]], align 4
759 // CHECK-NOUSE-I386-NEXT: [[TMP81:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 7
760 // CHECK-NOUSE-I386-NEXT: store i8** null, i8*** [[TMP81]], align 4
761 // CHECK-NOUSE-I386-NEXT: [[TMP82:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]], i32 0, i32 8
762 // CHECK-NOUSE-I386-NEXT: store i64 0, i64* [[TMP82]], align 8
763 // CHECK-NOUSE-I386-NEXT: [[TMP83:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS6]])
764 // CHECK-NOUSE-I386-NEXT: [[TMP84:%.*]] = icmp ne i32 [[TMP83]], 0
765 // CHECK-NOUSE-I386-NEXT: br i1 [[TMP84]], label [[OMP_OFFLOAD_FAILED7:%.*]], label [[OMP_OFFLOAD_CONT8:%.*]]
766 // CHECK-NOUSE-I386: omp_offload.failed7:
767 // CHECK-NOUSE-I386-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__Z20explicit_maps_singlei_l796() #[[ATTR3]]
768 // CHECK-NOUSE-I386-NEXT: br label [[OMP_OFFLOAD_CONT8]]
769 // CHECK-NOUSE-I386: omp_offload.cont8:
770 // CHECK-NOUSE-I386-NEXT: ret void
771 //
explicit_maps_single(int ii)772 void explicit_maps_single(int ii) {
773
774 // Map of a scalar.
775 int a = ii;
776
777 struct ST st1;
778 struct ST st2;
779
780 // Make sure the struct picks up ompx_hold even if another element of the
781 // struct doesn't have ompx_hold.
782 #pragma omp target map(tofrom : st1.i) \
783 map(ompx_hold, tofrom : a, st1.j, st2.i) \
784 map(tofrom : st2.j)
785 {
786 #ifdef USE
787 st1.i++;
788 a++;
789 st1.j++;
790 st2.i++;
791 st2.j++;
792 #endif
793 }
794
795 // Always Close Hold.
796 #pragma omp target map(always close ompx_hold tofrom: a)
797 {
798 #ifdef USE
799 a++;
800 #endif
801 }
802 }
803
804 // CHECK-USE-PPC64LE-LABEL: @_ZN2ST20test_present_membersEv(
805 // CHECK-USE-PPC64LE-NEXT: entry:
806 // CHECK-USE-PPC64LE-NEXT: [[THIS_ADDR:%.*]] = alloca %struct.ST*, align 8
807 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x i8*], align 8
808 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x i8*], align 8
809 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x i8*], align 8
810 // CHECK-USE-PPC64LE-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [3 x i64], align 8
811 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[THIS:%.*]], %struct.ST** [[THIS_ADDR]], align 8
812 // CHECK-USE-PPC64LE-NEXT: [[THIS1:%.*]] = load %struct.ST*, %struct.ST** [[THIS_ADDR]], align 8
813 // CHECK-USE-PPC64LE-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST:%.*]], %struct.ST* [[THIS1]], i32 0, i32 0
814 // CHECK-USE-PPC64LE-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[THIS1]], i32 0, i32 1
815 // CHECK-USE-PPC64LE-NEXT: [[TMP0:%.*]] = getelementptr i32, i32* [[J]], i32 1
816 // CHECK-USE-PPC64LE-NEXT: [[TMP1:%.*]] = bitcast i32* [[I]] to i8*
817 // CHECK-USE-PPC64LE-NEXT: [[TMP2:%.*]] = bitcast i32* [[TMP0]] to i8*
818 // CHECK-USE-PPC64LE-NEXT: [[TMP3:%.*]] = ptrtoint i8* [[TMP2]] to i64
819 // CHECK-USE-PPC64LE-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP1]] to i64
820 // CHECK-USE-PPC64LE-NEXT: [[TMP5:%.*]] = sub i64 [[TMP3]], [[TMP4]]
821 // CHECK-USE-PPC64LE-NEXT: [[TMP6:%.*]] = sdiv exact i64 [[TMP5]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
822 // CHECK-USE-PPC64LE-NEXT: [[TMP7:%.*]] = bitcast [3 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
823 // CHECK-USE-PPC64LE-NEXT: call void @llvm.memcpy.p0i8.p0i8.i64(i8* align 8 [[TMP7]], i8* align 8 bitcast ([3 x i64]* @.offload_sizes.3 to i8*), i64 24, i1 false)
824 // CHECK-USE-PPC64LE-NEXT: [[TMP8:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
825 // CHECK-USE-PPC64LE-NEXT: [[TMP9:%.*]] = bitcast i8** [[TMP8]] to %struct.ST**
826 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP9]], align 8
827 // CHECK-USE-PPC64LE-NEXT: [[TMP10:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
828 // CHECK-USE-PPC64LE-NEXT: [[TMP11:%.*]] = bitcast i8** [[TMP10]] to i32**
829 // CHECK-USE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP11]], align 8
830 // CHECK-USE-PPC64LE-NEXT: [[TMP12:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
831 // CHECK-USE-PPC64LE-NEXT: store i64 [[TMP6]], i64* [[TMP12]], align 8
832 // CHECK-USE-PPC64LE-NEXT: [[TMP13:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0
833 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP13]], align 8
834 // CHECK-USE-PPC64LE-NEXT: [[TMP14:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
835 // CHECK-USE-PPC64LE-NEXT: [[TMP15:%.*]] = bitcast i8** [[TMP14]] to %struct.ST**
836 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP15]], align 8
837 // CHECK-USE-PPC64LE-NEXT: [[TMP16:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
838 // CHECK-USE-PPC64LE-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to i32**
839 // CHECK-USE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP17]], align 8
840 // CHECK-USE-PPC64LE-NEXT: [[TMP18:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1
841 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP18]], align 8
842 // CHECK-USE-PPC64LE-NEXT: [[TMP19:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
843 // CHECK-USE-PPC64LE-NEXT: [[TMP20:%.*]] = bitcast i8** [[TMP19]] to %struct.ST**
844 // CHECK-USE-PPC64LE-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP20]], align 8
845 // CHECK-USE-PPC64LE-NEXT: [[TMP21:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
846 // CHECK-USE-PPC64LE-NEXT: [[TMP22:%.*]] = bitcast i8** [[TMP21]] to i32**
847 // CHECK-USE-PPC64LE-NEXT: store i32* [[J]], i32** [[TMP22]], align 8
848 // CHECK-USE-PPC64LE-NEXT: [[TMP23:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2
849 // CHECK-USE-PPC64LE-NEXT: store i8* null, i8** [[TMP23]], align 8
850 // CHECK-USE-PPC64LE-NEXT: [[TMP24:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
851 // CHECK-USE-PPC64LE-NEXT: [[TMP25:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
852 // CHECK-USE-PPC64LE-NEXT: [[TMP26:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
853 // CHECK-USE-PPC64LE-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
854 // CHECK-USE-PPC64LE-NEXT: [[TMP27:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
855 // CHECK-USE-PPC64LE-NEXT: store i32 1, i32* [[TMP27]], align 4
856 // CHECK-USE-PPC64LE-NEXT: [[TMP28:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
857 // CHECK-USE-PPC64LE-NEXT: store i32 3, i32* [[TMP28]], align 4
858 // CHECK-USE-PPC64LE-NEXT: [[TMP29:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
859 // CHECK-USE-PPC64LE-NEXT: store i8** [[TMP24]], i8*** [[TMP29]], align 8
860 // CHECK-USE-PPC64LE-NEXT: [[TMP30:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
861 // CHECK-USE-PPC64LE-NEXT: store i8** [[TMP25]], i8*** [[TMP30]], align 8
862 // CHECK-USE-PPC64LE-NEXT: [[TMP31:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
863 // CHECK-USE-PPC64LE-NEXT: store i64* [[TMP26]], i64** [[TMP31]], align 8
864 // CHECK-USE-PPC64LE-NEXT: [[TMP32:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
865 // CHECK-USE-PPC64LE-NEXT: store i64* getelementptr inbounds ([3 x i64], [3 x i64]* @.offload_maptypes.4, i32 0, i32 0), i64** [[TMP32]], align 8
866 // CHECK-USE-PPC64LE-NEXT: [[TMP33:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
867 // CHECK-USE-PPC64LE-NEXT: store i8** null, i8*** [[TMP33]], align 8
868 // CHECK-USE-PPC64LE-NEXT: [[TMP34:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
869 // CHECK-USE-PPC64LE-NEXT: store i8** null, i8*** [[TMP34]], align 8
870 // CHECK-USE-PPC64LE-NEXT: [[TMP35:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
871 // CHECK-USE-PPC64LE-NEXT: store i64 0, i64* [[TMP35]], align 8
872 // CHECK-USE-PPC64LE-NEXT: [[TMP36:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
873 // CHECK-USE-PPC64LE-NEXT: [[TMP37:%.*]] = icmp ne i32 [[TMP36]], 0
874 // CHECK-USE-PPC64LE-NEXT: br i1 [[TMP37]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
875 // CHECK-USE-PPC64LE: omp_offload.failed:
876 // CHECK-USE-PPC64LE-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115(%struct.ST* [[THIS1]]) #[[ATTR3]]
877 // CHECK-USE-PPC64LE-NEXT: br label [[OMP_OFFLOAD_CONT]]
878 // CHECK-USE-PPC64LE: omp_offload.cont:
879 // CHECK-USE-PPC64LE-NEXT: ret void
880 //
881 // CHECK-USE-I386-LABEL: @_ZN2ST20test_present_membersEv(
882 // CHECK-USE-I386-NEXT: entry:
883 // CHECK-USE-I386-NEXT: [[THIS_ADDR:%.*]] = alloca %struct.ST*, align 4
884 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x i8*], align 4
885 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x i8*], align 4
886 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x i8*], align 4
887 // CHECK-USE-I386-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [3 x i64], align 4
888 // CHECK-USE-I386-NEXT: store %struct.ST* [[THIS:%.*]], %struct.ST** [[THIS_ADDR]], align 4
889 // CHECK-USE-I386-NEXT: [[THIS1:%.*]] = load %struct.ST*, %struct.ST** [[THIS_ADDR]], align 4
890 // CHECK-USE-I386-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST:%.*]], %struct.ST* [[THIS1]], i32 0, i32 0
891 // CHECK-USE-I386-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[THIS1]], i32 0, i32 1
892 // CHECK-USE-I386-NEXT: [[TMP0:%.*]] = getelementptr i32, i32* [[J]], i32 1
893 // CHECK-USE-I386-NEXT: [[TMP1:%.*]] = bitcast i32* [[I]] to i8*
894 // CHECK-USE-I386-NEXT: [[TMP2:%.*]] = bitcast i32* [[TMP0]] to i8*
895 // CHECK-USE-I386-NEXT: [[TMP3:%.*]] = ptrtoint i8* [[TMP2]] to i64
896 // CHECK-USE-I386-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP1]] to i64
897 // CHECK-USE-I386-NEXT: [[TMP5:%.*]] = sub i64 [[TMP3]], [[TMP4]]
898 // CHECK-USE-I386-NEXT: [[TMP6:%.*]] = sdiv exact i64 [[TMP5]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
899 // CHECK-USE-I386-NEXT: [[TMP7:%.*]] = bitcast [3 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
900 // CHECK-USE-I386-NEXT: call void @llvm.memcpy.p0i8.p0i8.i32(i8* align 4 [[TMP7]], i8* align 4 bitcast ([3 x i64]* @.offload_sizes.3 to i8*), i32 24, i1 false)
901 // CHECK-USE-I386-NEXT: [[TMP8:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
902 // CHECK-USE-I386-NEXT: [[TMP9:%.*]] = bitcast i8** [[TMP8]] to %struct.ST**
903 // CHECK-USE-I386-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP9]], align 4
904 // CHECK-USE-I386-NEXT: [[TMP10:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
905 // CHECK-USE-I386-NEXT: [[TMP11:%.*]] = bitcast i8** [[TMP10]] to i32**
906 // CHECK-USE-I386-NEXT: store i32* [[I]], i32** [[TMP11]], align 4
907 // CHECK-USE-I386-NEXT: [[TMP12:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
908 // CHECK-USE-I386-NEXT: store i64 [[TMP6]], i64* [[TMP12]], align 4
909 // CHECK-USE-I386-NEXT: [[TMP13:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0
910 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP13]], align 4
911 // CHECK-USE-I386-NEXT: [[TMP14:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
912 // CHECK-USE-I386-NEXT: [[TMP15:%.*]] = bitcast i8** [[TMP14]] to %struct.ST**
913 // CHECK-USE-I386-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP15]], align 4
914 // CHECK-USE-I386-NEXT: [[TMP16:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
915 // CHECK-USE-I386-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to i32**
916 // CHECK-USE-I386-NEXT: store i32* [[I]], i32** [[TMP17]], align 4
917 // CHECK-USE-I386-NEXT: [[TMP18:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 1
918 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP18]], align 4
919 // CHECK-USE-I386-NEXT: [[TMP19:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
920 // CHECK-USE-I386-NEXT: [[TMP20:%.*]] = bitcast i8** [[TMP19]] to %struct.ST**
921 // CHECK-USE-I386-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP20]], align 4
922 // CHECK-USE-I386-NEXT: [[TMP21:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
923 // CHECK-USE-I386-NEXT: [[TMP22:%.*]] = bitcast i8** [[TMP21]] to i32**
924 // CHECK-USE-I386-NEXT: store i32* [[J]], i32** [[TMP22]], align 4
925 // CHECK-USE-I386-NEXT: [[TMP23:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 2
926 // CHECK-USE-I386-NEXT: store i8* null, i8** [[TMP23]], align 4
927 // CHECK-USE-I386-NEXT: [[TMP24:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
928 // CHECK-USE-I386-NEXT: [[TMP25:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
929 // CHECK-USE-I386-NEXT: [[TMP26:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
930 // CHECK-USE-I386-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
931 // CHECK-USE-I386-NEXT: [[TMP27:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
932 // CHECK-USE-I386-NEXT: store i32 1, i32* [[TMP27]], align 4
933 // CHECK-USE-I386-NEXT: [[TMP28:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
934 // CHECK-USE-I386-NEXT: store i32 3, i32* [[TMP28]], align 4
935 // CHECK-USE-I386-NEXT: [[TMP29:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
936 // CHECK-USE-I386-NEXT: store i8** [[TMP24]], i8*** [[TMP29]], align 4
937 // CHECK-USE-I386-NEXT: [[TMP30:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
938 // CHECK-USE-I386-NEXT: store i8** [[TMP25]], i8*** [[TMP30]], align 4
939 // CHECK-USE-I386-NEXT: [[TMP31:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
940 // CHECK-USE-I386-NEXT: store i64* [[TMP26]], i64** [[TMP31]], align 4
941 // CHECK-USE-I386-NEXT: [[TMP32:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
942 // CHECK-USE-I386-NEXT: store i64* getelementptr inbounds ([3 x i64], [3 x i64]* @.offload_maptypes.4, i32 0, i32 0), i64** [[TMP32]], align 4
943 // CHECK-USE-I386-NEXT: [[TMP33:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
944 // CHECK-USE-I386-NEXT: store i8** null, i8*** [[TMP33]], align 4
945 // CHECK-USE-I386-NEXT: [[TMP34:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
946 // CHECK-USE-I386-NEXT: store i8** null, i8*** [[TMP34]], align 4
947 // CHECK-USE-I386-NEXT: [[TMP35:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
948 // CHECK-USE-I386-NEXT: store i64 0, i64* [[TMP35]], align 8
949 // CHECK-USE-I386-NEXT: [[TMP36:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
950 // CHECK-USE-I386-NEXT: [[TMP37:%.*]] = icmp ne i32 [[TMP36]], 0
951 // CHECK-USE-I386-NEXT: br i1 [[TMP37]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
952 // CHECK-USE-I386: omp_offload.failed:
953 // CHECK-USE-I386-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115(%struct.ST* [[THIS1]]) #[[ATTR3]]
954 // CHECK-USE-I386-NEXT: br label [[OMP_OFFLOAD_CONT]]
955 // CHECK-USE-I386: omp_offload.cont:
956 // CHECK-USE-I386-NEXT: ret void
957 //
958 // CHECK-NOUSE-PPC64LE-LABEL: @_ZN2ST20test_present_membersEv(
959 // CHECK-NOUSE-PPC64LE-NEXT: entry:
960 // CHECK-NOUSE-PPC64LE-NEXT: [[THIS_ADDR:%.*]] = alloca %struct.ST*, align 8
961 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x i8*], align 8
962 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x i8*], align 8
963 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x i8*], align 8
964 // CHECK-NOUSE-PPC64LE-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [3 x i64], align 8
965 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[THIS:%.*]], %struct.ST** [[THIS_ADDR]], align 8
966 // CHECK-NOUSE-PPC64LE-NEXT: [[THIS1:%.*]] = load %struct.ST*, %struct.ST** [[THIS_ADDR]], align 8
967 // CHECK-NOUSE-PPC64LE-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST:%.*]], %struct.ST* [[THIS1]], i32 0, i32 0
968 // CHECK-NOUSE-PPC64LE-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[THIS1]], i32 0, i32 1
969 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP0:%.*]] = getelementptr i32, i32* [[J]], i32 1
970 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP1:%.*]] = bitcast i32* [[I]] to i8*
971 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP2:%.*]] = bitcast i32* [[TMP0]] to i8*
972 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP3:%.*]] = ptrtoint i8* [[TMP2]] to i64
973 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP1]] to i64
974 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP5:%.*]] = sub i64 [[TMP3]], [[TMP4]]
975 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP6:%.*]] = sdiv exact i64 [[TMP5]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
976 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP7:%.*]] = bitcast [3 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
977 // CHECK-NOUSE-PPC64LE-NEXT: call void @llvm.memcpy.p0i8.p0i8.i64(i8* align 8 [[TMP7]], i8* align 8 bitcast ([3 x i64]* @.offload_sizes.3 to i8*), i64 24, i1 false)
978 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP8:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
979 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP9:%.*]] = bitcast i8** [[TMP8]] to %struct.ST**
980 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP9]], align 8
981 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP10:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
982 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP11:%.*]] = bitcast i8** [[TMP10]] to i32**
983 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP11]], align 8
984 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP12:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
985 // CHECK-NOUSE-PPC64LE-NEXT: store i64 [[TMP6]], i64* [[TMP12]], align 8
986 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP13:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 0
987 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP13]], align 8
988 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP14:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
989 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP15:%.*]] = bitcast i8** [[TMP14]] to %struct.ST**
990 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP15]], align 8
991 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP16:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
992 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to i32**
993 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[I]], i32** [[TMP17]], align 8
994 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP18:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 1
995 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP18]], align 8
996 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP19:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
997 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP20:%.*]] = bitcast i8** [[TMP19]] to %struct.ST**
998 // CHECK-NOUSE-PPC64LE-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP20]], align 8
999 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP21:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
1000 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP22:%.*]] = bitcast i8** [[TMP21]] to i32**
1001 // CHECK-NOUSE-PPC64LE-NEXT: store i32* [[J]], i32** [[TMP22]], align 8
1002 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP23:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i64 0, i64 2
1003 // CHECK-NOUSE-PPC64LE-NEXT: store i8* null, i8** [[TMP23]], align 8
1004 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP24:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
1005 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP25:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
1006 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP26:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
1007 // CHECK-NOUSE-PPC64LE-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
1008 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP27:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
1009 // CHECK-NOUSE-PPC64LE-NEXT: store i32 1, i32* [[TMP27]], align 4
1010 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP28:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
1011 // CHECK-NOUSE-PPC64LE-NEXT: store i32 3, i32* [[TMP28]], align 4
1012 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP29:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
1013 // CHECK-NOUSE-PPC64LE-NEXT: store i8** [[TMP24]], i8*** [[TMP29]], align 8
1014 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP30:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
1015 // CHECK-NOUSE-PPC64LE-NEXT: store i8** [[TMP25]], i8*** [[TMP30]], align 8
1016 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP31:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
1017 // CHECK-NOUSE-PPC64LE-NEXT: store i64* [[TMP26]], i64** [[TMP31]], align 8
1018 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP32:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
1019 // CHECK-NOUSE-PPC64LE-NEXT: store i64* getelementptr inbounds ([3 x i64], [3 x i64]* @.offload_maptypes.4, i32 0, i32 0), i64** [[TMP32]], align 8
1020 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP33:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
1021 // CHECK-NOUSE-PPC64LE-NEXT: store i8** null, i8*** [[TMP33]], align 8
1022 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP34:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
1023 // CHECK-NOUSE-PPC64LE-NEXT: store i8** null, i8*** [[TMP34]], align 8
1024 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP35:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
1025 // CHECK-NOUSE-PPC64LE-NEXT: store i64 0, i64* [[TMP35]], align 8
1026 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP36:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
1027 // CHECK-NOUSE-PPC64LE-NEXT: [[TMP37:%.*]] = icmp ne i32 [[TMP36]], 0
1028 // CHECK-NOUSE-PPC64LE-NEXT: br i1 [[TMP37]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
1029 // CHECK-NOUSE-PPC64LE: omp_offload.failed:
1030 // CHECK-NOUSE-PPC64LE-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115() #[[ATTR3]]
1031 // CHECK-NOUSE-PPC64LE-NEXT: br label [[OMP_OFFLOAD_CONT]]
1032 // CHECK-NOUSE-PPC64LE: omp_offload.cont:
1033 // CHECK-NOUSE-PPC64LE-NEXT: ret void
1034 //
1035 // CHECK-NOUSE-I386-LABEL: @_ZN2ST20test_present_membersEv(
1036 // CHECK-NOUSE-I386-NEXT: entry:
1037 // CHECK-NOUSE-I386-NEXT: [[THIS_ADDR:%.*]] = alloca %struct.ST*, align 4
1038 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_BASEPTRS:%.*]] = alloca [3 x i8*], align 4
1039 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_PTRS:%.*]] = alloca [3 x i8*], align 4
1040 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_MAPPERS:%.*]] = alloca [3 x i8*], align 4
1041 // CHECK-NOUSE-I386-NEXT: [[DOTOFFLOAD_SIZES:%.*]] = alloca [3 x i64], align 4
1042 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[THIS:%.*]], %struct.ST** [[THIS_ADDR]], align 4
1043 // CHECK-NOUSE-I386-NEXT: [[THIS1:%.*]] = load %struct.ST*, %struct.ST** [[THIS_ADDR]], align 4
1044 // CHECK-NOUSE-I386-NEXT: [[I:%.*]] = getelementptr inbounds [[STRUCT_ST:%.*]], %struct.ST* [[THIS1]], i32 0, i32 0
1045 // CHECK-NOUSE-I386-NEXT: [[J:%.*]] = getelementptr inbounds [[STRUCT_ST]], %struct.ST* [[THIS1]], i32 0, i32 1
1046 // CHECK-NOUSE-I386-NEXT: [[TMP0:%.*]] = getelementptr i32, i32* [[J]], i32 1
1047 // CHECK-NOUSE-I386-NEXT: [[TMP1:%.*]] = bitcast i32* [[I]] to i8*
1048 // CHECK-NOUSE-I386-NEXT: [[TMP2:%.*]] = bitcast i32* [[TMP0]] to i8*
1049 // CHECK-NOUSE-I386-NEXT: [[TMP3:%.*]] = ptrtoint i8* [[TMP2]] to i64
1050 // CHECK-NOUSE-I386-NEXT: [[TMP4:%.*]] = ptrtoint i8* [[TMP1]] to i64
1051 // CHECK-NOUSE-I386-NEXT: [[TMP5:%.*]] = sub i64 [[TMP3]], [[TMP4]]
1052 // CHECK-NOUSE-I386-NEXT: [[TMP6:%.*]] = sdiv exact i64 [[TMP5]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64)
1053 // CHECK-NOUSE-I386-NEXT: [[TMP7:%.*]] = bitcast [3 x i64]* [[DOTOFFLOAD_SIZES]] to i8*
1054 // CHECK-NOUSE-I386-NEXT: call void @llvm.memcpy.p0i8.p0i8.i32(i8* align 4 [[TMP7]], i8* align 4 bitcast ([3 x i64]* @.offload_sizes.3 to i8*), i32 24, i1 false)
1055 // CHECK-NOUSE-I386-NEXT: [[TMP8:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
1056 // CHECK-NOUSE-I386-NEXT: [[TMP9:%.*]] = bitcast i8** [[TMP8]] to %struct.ST**
1057 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP9]], align 4
1058 // CHECK-NOUSE-I386-NEXT: [[TMP10:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
1059 // CHECK-NOUSE-I386-NEXT: [[TMP11:%.*]] = bitcast i8** [[TMP10]] to i32**
1060 // CHECK-NOUSE-I386-NEXT: store i32* [[I]], i32** [[TMP11]], align 4
1061 // CHECK-NOUSE-I386-NEXT: [[TMP12:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
1062 // CHECK-NOUSE-I386-NEXT: store i64 [[TMP6]], i64* [[TMP12]], align 4
1063 // CHECK-NOUSE-I386-NEXT: [[TMP13:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 0
1064 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP13]], align 4
1065 // CHECK-NOUSE-I386-NEXT: [[TMP14:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 1
1066 // CHECK-NOUSE-I386-NEXT: [[TMP15:%.*]] = bitcast i8** [[TMP14]] to %struct.ST**
1067 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP15]], align 4
1068 // CHECK-NOUSE-I386-NEXT: [[TMP16:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 1
1069 // CHECK-NOUSE-I386-NEXT: [[TMP17:%.*]] = bitcast i8** [[TMP16]] to i32**
1070 // CHECK-NOUSE-I386-NEXT: store i32* [[I]], i32** [[TMP17]], align 4
1071 // CHECK-NOUSE-I386-NEXT: [[TMP18:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 1
1072 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP18]], align 4
1073 // CHECK-NOUSE-I386-NEXT: [[TMP19:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 2
1074 // CHECK-NOUSE-I386-NEXT: [[TMP20:%.*]] = bitcast i8** [[TMP19]] to %struct.ST**
1075 // CHECK-NOUSE-I386-NEXT: store %struct.ST* [[THIS1]], %struct.ST** [[TMP20]], align 4
1076 // CHECK-NOUSE-I386-NEXT: [[TMP21:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 2
1077 // CHECK-NOUSE-I386-NEXT: [[TMP22:%.*]] = bitcast i8** [[TMP21]] to i32**
1078 // CHECK-NOUSE-I386-NEXT: store i32* [[J]], i32** [[TMP22]], align 4
1079 // CHECK-NOUSE-I386-NEXT: [[TMP23:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_MAPPERS]], i32 0, i32 2
1080 // CHECK-NOUSE-I386-NEXT: store i8* null, i8** [[TMP23]], align 4
1081 // CHECK-NOUSE-I386-NEXT: [[TMP24:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_BASEPTRS]], i32 0, i32 0
1082 // CHECK-NOUSE-I386-NEXT: [[TMP25:%.*]] = getelementptr inbounds [3 x i8*], [3 x i8*]* [[DOTOFFLOAD_PTRS]], i32 0, i32 0
1083 // CHECK-NOUSE-I386-NEXT: [[TMP26:%.*]] = getelementptr inbounds [3 x i64], [3 x i64]* [[DOTOFFLOAD_SIZES]], i32 0, i32 0
1084 // CHECK-NOUSE-I386-NEXT: [[KERNEL_ARGS:%.*]] = alloca [[STRUCT___TGT_KERNEL_ARGUMENTS:%.*]], align 8
1085 // CHECK-NOUSE-I386-NEXT: [[TMP27:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 0
1086 // CHECK-NOUSE-I386-NEXT: store i32 1, i32* [[TMP27]], align 4
1087 // CHECK-NOUSE-I386-NEXT: [[TMP28:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 1
1088 // CHECK-NOUSE-I386-NEXT: store i32 3, i32* [[TMP28]], align 4
1089 // CHECK-NOUSE-I386-NEXT: [[TMP29:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 2
1090 // CHECK-NOUSE-I386-NEXT: store i8** [[TMP24]], i8*** [[TMP29]], align 4
1091 // CHECK-NOUSE-I386-NEXT: [[TMP30:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 3
1092 // CHECK-NOUSE-I386-NEXT: store i8** [[TMP25]], i8*** [[TMP30]], align 4
1093 // CHECK-NOUSE-I386-NEXT: [[TMP31:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 4
1094 // CHECK-NOUSE-I386-NEXT: store i64* [[TMP26]], i64** [[TMP31]], align 4
1095 // CHECK-NOUSE-I386-NEXT: [[TMP32:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 5
1096 // CHECK-NOUSE-I386-NEXT: store i64* getelementptr inbounds ([3 x i64], [3 x i64]* @.offload_maptypes.4, i32 0, i32 0), i64** [[TMP32]], align 4
1097 // CHECK-NOUSE-I386-NEXT: [[TMP33:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 6
1098 // CHECK-NOUSE-I386-NEXT: store i8** null, i8*** [[TMP33]], align 4
1099 // CHECK-NOUSE-I386-NEXT: [[TMP34:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 7
1100 // CHECK-NOUSE-I386-NEXT: store i8** null, i8*** [[TMP34]], align 4
1101 // CHECK-NOUSE-I386-NEXT: [[TMP35:%.*]] = getelementptr inbounds [[STRUCT___TGT_KERNEL_ARGUMENTS]], %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]], i32 0, i32 8
1102 // CHECK-NOUSE-I386-NEXT: store i64 0, i64* [[TMP35]], align 8
1103 // CHECK-NOUSE-I386-NEXT: [[TMP36:%.*]] = call i32 @__tgt_target_kernel(%struct.ident_t* @[[GLOB1]], i64 -1, i32 -1, i32 0, i8* @.{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115.region_id, %struct.__tgt_kernel_arguments* [[KERNEL_ARGS]])
1104 // CHECK-NOUSE-I386-NEXT: [[TMP37:%.*]] = icmp ne i32 [[TMP36]], 0
1105 // CHECK-NOUSE-I386-NEXT: br i1 [[TMP37]], label [[OMP_OFFLOAD_FAILED:%.*]], label [[OMP_OFFLOAD_CONT:%.*]]
1106 // CHECK-NOUSE-I386: omp_offload.failed:
1107 // CHECK-NOUSE-I386-NEXT: call void @{{__omp_offloading_[0-9a-z]+_[0-9a-z]+}}__ZN2ST20test_present_membersEv_l1115() #[[ATTR3]]
1108 // CHECK-NOUSE-I386-NEXT: br label [[OMP_OFFLOAD_CONT]]
1109 // CHECK-NOUSE-I386: omp_offload.cont:
1110 // CHECK-NOUSE-I386-NEXT: ret void
1111 //
test_present_members()1112 void ST::test_present_members() {
1113 // Make sure the struct picks up ompx_hold even if another element of the
1114 // struct doesn't have ompx_hold.
1115 #pragma omp target map(tofrom : i) map(ompx_hold, tofrom : j)
1116 {
1117 #ifdef USE
1118 i++;
1119 j++;
1120 #endif
1121 }
1122 }
1123
1124 #endif
1125