1 // SIMD-ONLY0-NOT: {{__kmpc|__tgt}} 2 3 // expected-no-diagnostics 4 #ifndef HEADER 5 #define HEADER 6 7 ///==========================================================================/// 8 // RUN: %clang_cc1 -DCK0 -verify -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK0 --check-prefix CK0-64 %s 9 // RUN: %clang_cc1 -DCK0 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 10 // RUN: %clang_cc1 -DCK0 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK0 --check-prefix CK0-64 %s 11 // RUN: %clang_cc1 -DCK0 -verify -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK0 --check-prefix CK0-32 %s 12 // RUN: %clang_cc1 -DCK0 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 13 // RUN: %clang_cc1 -DCK0 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK0 --check-prefix CK0-32 %s 14 15 // RUN: %clang_cc1 -DCK0 -verify -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 16 // RUN: %clang_cc1 -DCK0 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 17 // RUN: %clang_cc1 -DCK0 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 18 // RUN: %clang_cc1 -DCK0 -verify -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 19 // RUN: %clang_cc1 -DCK0 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 20 // RUN: %clang_cc1 -DCK0 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 21 22 #ifdef CK0 23 // Mapper function code generation and runtime interface. 24 25 // CK0: [[IDENT_T:%.+]] = type { i32, i32, i32, i32, i8* } 26 // CK0: [[ENTRY:%.+]] = type { i8*, i8*, i[[SZ:32|64]], i32, i32 } 27 // CK0: [[ANON_T:%.+]] = type { %class.C* } 28 // CK0: [[ANON_T_0:%.+]] = type { %class.C* } 29 // CK0: [[KMP_TASK_T_WITH_PRIVATES:%.+]] = type { [[KMP_TASK_T:%[^,]+]], [[KMP_PRIVATES_T:%.+]] } 30 // CK0: [[KMP_TASK_T]] = type { i8*, i32 (i32, i8*)*, i32, %{{[^,]+}}, %{{[^,]+}} } 31 // CK0-32: [[KMP_PRIVATES_T]] = type { [1 x i64], [1 x i8*], [1 x i8*], [1 x i8*] } 32 // CK0-64: [[KMP_PRIVATES_T]] = type { [1 x i8*], [1 x i8*], [1 x i64], [1 x i8*] } 33 // CK0: [[KMP_TASK_T_WITH_PRIVATES_1:%.+]] = type { [[KMP_TASK_T]], [[KMP_PRIVATES_T_2:%.+]] } 34 // CK0-32: [[KMP_PRIVATES_T_2]] = type { [1 x i64], [1 x i8*], [1 x i8*], [1 x i8*] } 35 // CK0-64: [[KMP_PRIVATES_T_2]] = type { [1 x i8*], [1 x i8*], [1 x i64], [1 x i8*] } 36 37 // CK0-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}.region_id = weak constant i8 0 38 // CK0-64: [[SIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 39 // CK0-32: [[SIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 40 // CK0: [[TYPES:@.+]] = {{.+}}constant [1 x i64] [i64 35] 41 // CK0-64: [[NWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 42 // CK0-32: [[NWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 43 // CK0: [[NWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 35] 44 // CK0-64: [[TEAMSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 45 // CK0-32: [[TEAMSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 46 // CK0: [[TEAMTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33] 47 // CK0-64: [[TEAMNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 48 // CK0-32: [[TEAMNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 49 // CK0: [[TEAMNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33] 50 // CK0-64: [[EDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 51 // CK0-32: [[EDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 52 // CK0: [[EDTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33] 53 // CK0-64: [[EDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 54 // CK0-32: [[EDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 55 // CK0: [[EDNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33] 56 // CK0-64: [[EXDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 57 // CK0-32: [[EXDSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 58 // CK0: [[EXDTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34] 59 // CK0-64: [[EXDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 60 // CK0-32: [[EXDNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 61 // CK0: [[EXDNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34] 62 // CK0-64: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 63 // CK0-32: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 64 // CK0: [[TTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 33] 65 // CK0-64: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 66 // CK0-32: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 67 // CK0: [[FTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34] 68 // CK0-64: [[FNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 69 // CK0-32: [[FNWSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 70 // CK0: [[FNWTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 34] 71 72 class C { 73 public: 74 int a; 75 double *b; 76 }; 77 78 #pragma omp declare mapper(id: C s) map(s.a, s.b[0:2]) 79 80 // CK0: define {{.*}}void [[MPRFUNC:@[.]omp_mapper[.].*C[.]id]](i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}}) 81 // CK0: store i8* %{{[^,]+}}, i8** [[HANDLEADDR:%[^,]+]] 82 // CK0: store i8* %{{[^,]+}}, i8** [[BPTRADDR:%[^,]+]] 83 // CK0: store i8* %{{[^,]+}}, i8** [[VPTRADDR:%[^,]+]] 84 // CK0: store i64 %{{[^,]+}}, i{{64|32}}* [[SIZEADDR:%[^,]+]] 85 // CK0: store i64 %{{[^,]+}}, i64* [[TYPEADDR:%[^,]+]] 86 // CK0-DAG: [[BYTESIZE:%.+]] = load i64, i64* [[SIZEADDR]] 87 // CK0-64-DAG: [[SIZE:%.+]] = udiv exact i64 [[BYTESIZE]], 16 88 // CK0-32-DAG: [[SIZE:%.+]] = udiv exact i64 [[BYTESIZE]], 8 89 // CK0-DAG: [[TYPE:%.+]] = load i64, i64* [[TYPEADDR]] 90 // CK0-DAG: [[HANDLE:%.+]] = load i8*, i8** [[HANDLEADDR]] 91 // CK0-DAG: [[PTRBEGIN:%.+]] = bitcast i8** [[VPTRADDR]] to %class.C** 92 // CK0-DAG: [[PTREND:%.+]] = getelementptr %class.C*, %class.C** [[PTRBEGIN]], i64 [[SIZE]] 93 // CK0-DAG: [[BPTR:%.+]] = load i8*, i8** [[BPTRADDR]] 94 // CK0-DAG: [[BEGIN:%.+]] = load i8*, i8** [[VPTRADDR]] 95 // CK0: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 96 // CK0: br i1 [[ISARRAY]], label %[[INITEVALDEL:[^,]+]], label %[[LHEAD:[^,]+]] 97 98 // CK0: [[INITEVALDEL]] 99 // CK0: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 100 // CK0: [[ISNOTDEL:%.+]] = icmp eq i64 [[TYPEDEL]], 0 101 // CK0: br i1 [[ISNOTDEL]], label %[[INIT:[^,]+]], label %[[LHEAD:[^,]+]] 102 // CK0: [[INIT]] 103 // CK0-64-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 16 104 // CK0-32-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 8 105 // CK0-DAG: [[ITYPE:%.+]] = and i64 [[TYPE]], -4 106 // CK0: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[ITYPE]]) 107 // CK0: br label %[[LHEAD:[^,]+]] 108 109 // CK0: [[LHEAD]] 110 // CK0: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]] 111 // CK0: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]] 112 // CK0: [[LBODY]] 113 // CK0: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ] 114 // CK0: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]] 115 // CK0-DAG: [[ABEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 0 116 // CK0-DAG: [[BBEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 1 117 // CK0-DAG: [[BBEGIN2:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 1 118 // CK0-DAG: [[BARRBEGIN:%.+]] = load double*, double** [[BBEGIN2]] 119 // CK0-DAG: [[BARRBEGINGEP:%.+]] = getelementptr inbounds double, double* [[BARRBEGIN]], i[[sz:64|32]] 0 120 // CK0-DAG: [[BEND:%.+]] = getelementptr double*, double** [[BBEGIN]], i32 1 121 // CK0-DAG: [[ABEGINV:%.+]] = bitcast i32* [[ABEGIN]] to i8* 122 // CK0-DAG: [[BENDV:%.+]] = bitcast double** [[BEND]] to i8* 123 // CK0-DAG: [[ABEGINI:%.+]] = ptrtoint i8* [[ABEGINV]] to i64 124 // CK0-DAG: [[BENDI:%.+]] = ptrtoint i8* [[BENDV]] to i64 125 // CK0-DAG: [[CSIZE:%.+]] = sub i64 [[BENDI]], [[ABEGINI]] 126 // CK0-DAG: [[CUSIZE:%.+]] = sdiv exact i64 [[CSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64) 127 // CK0-DAG: [[BPTRADDR0BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 128 // CK0-DAG: [[PTRADDR0BC:%.+]] = bitcast i32* [[ABEGIN]] to i8* 129 // CK0-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]]) 130 // CK0-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48 131 // CK0-DAG: br label %[[MEMBER:[^,]+]] 132 // CK0-DAG: [[MEMBER]] 133 // CK0-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 134 // CK0-DAG: [[MEMBERCOM]] 135 // CK0-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]] 136 // CK0-DAG: br label %[[LTYPE]] 137 // CK0-DAG: [[LTYPE]] 138 // CK0-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 139 // CK0-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 140 // CK0-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 141 // CK0-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 142 // CK0-DAG: [[ALLOC]] 143 // CK0-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 144 // CK0-DAG: br label %[[TYEND:[^,]+]] 145 // CK0-DAG: [[ALLOCELSE]] 146 // CK0-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 147 // CK0-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 148 // CK0-DAG: [[TO]] 149 // CK0-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 150 // CK0-DAG: br label %[[TYEND]] 151 // CK0-DAG: [[TOELSE]] 152 // CK0-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 153 // CK0-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 154 // CK0-DAG: [[FROM]] 155 // CK0-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 156 // CK0-DAG: br label %[[TYEND]] 157 // CK0-DAG: [[TYEND]] 158 // CK0-DAG: [[PHITYPE0:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 159 // CK0: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR0BC]], i8* [[PTRADDR0BC]], i64 [[CUSIZE]], i64 [[PHITYPE0]]) 160 // CK0-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 161 // CK0-DAG: [[PTRADDR1BC:%.+]] = bitcast i32* [[ABEGIN]] to i8* 162 // CK0-DAG: br label %[[MEMBER:[^,]+]] 163 // CK0-DAG: [[MEMBER]] 164 // CK0-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 165 // CK0-DAG: [[MEMBERCOM]] 166 // 281474976710659 == 0x1,000,000,003 167 // CK0-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710659, [[SHIPRESIZE]] 168 // CK0-DAG: br label %[[LTYPE]] 169 // CK0-DAG: [[LTYPE]] 170 // CK0-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710659, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 171 // CK0-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 172 // CK0-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 173 // CK0-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 174 // CK0-DAG: [[ALLOC]] 175 // CK0-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 176 // CK0-DAG: br label %[[TYEND:[^,]+]] 177 // CK0-DAG: [[ALLOCELSE]] 178 // CK0-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 179 // CK0-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 180 // CK0-DAG: [[TO]] 181 // CK0-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 182 // CK0-DAG: br label %[[TYEND]] 183 // CK0-DAG: [[TOELSE]] 184 // CK0-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 185 // CK0-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 186 // CK0-DAG: [[FROM]] 187 // CK0-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 188 // CK0-DAG: br label %[[TYEND]] 189 // CK0-DAG: [[TYEND]] 190 // CK0-DAG: [[TYPE1:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 191 // CK0: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR1BC]], i8* [[PTRADDR1BC]], i64 4, i64 [[TYPE1]]) 192 // CK0-DAG: [[BPTRADDR2BC:%.+]] = bitcast double** [[BBEGIN]] to i8* 193 // CK0-DAG: [[PTRADDR2BC:%.+]] = bitcast double* [[BARRBEGINGEP]] to i8* 194 // CK0-DAG: br label %[[MEMBER:[^,]+]] 195 // CK0-DAG: [[MEMBER]] 196 // CK0-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 197 // CK0-DAG: [[MEMBERCOM]] 198 // 281474976710675 == 0x1,000,000,013 199 // CK0-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710675, [[SHIPRESIZE]] 200 // CK0-DAG: br label %[[LTYPE]] 201 // CK0-DAG: [[LTYPE]] 202 // CK0-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710675, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 203 // CK0-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 204 // CK0-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 205 // CK0-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 206 // CK0-DAG: [[ALLOC]] 207 // CK0-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 208 // CK0-DAG: br label %[[TYEND:[^,]+]] 209 // CK0-DAG: [[ALLOCELSE]] 210 // CK0-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 211 // CK0-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 212 // CK0-DAG: [[TO]] 213 // CK0-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 214 // CK0-DAG: br label %[[TYEND]] 215 // CK0-DAG: [[TOELSE]] 216 // CK0-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 217 // CK0-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 218 // CK0-DAG: [[FROM]] 219 // CK0-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 220 // CK0-DAG: br label %[[TYEND]] 221 // CK0-DAG: [[TYEND]] 222 // CK0-DAG: [[TYPE2:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 223 // CK0: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR2BC]], i8* [[PTRADDR2BC]], i64 16, i64 [[TYPE2]]) 224 // CK0: [[PTRNEXT]] = getelementptr %class.C*, %class.C** [[PTR]], i32 1 225 // CK0: [[ISDONE:%.+]] = icmp eq %class.C** [[PTRNEXT]], [[PTREND]] 226 // CK0: br i1 [[ISDONE]], label %[[LEXIT:[^,]+]], label %[[LBODY]] 227 228 // CK0: [[LEXIT]] 229 // CK0: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 230 // CK0: br i1 [[ISARRAY]], label %[[EVALDEL:[^,]+]], label %[[DONE]] 231 // CK0: [[EVALDEL]] 232 // CK0: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 233 // CK0: [[ISDEL:%.+]] = icmp ne i64 [[TYPEDEL]], 0 234 // CK0: br i1 [[ISDEL]], label %[[DEL:[^,]+]], label %[[DONE]] 235 // CK0: [[DEL]] 236 // CK0-64-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 16 237 // CK0-32-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 8 238 // CK0-DAG: [[DTYPE:%.+]] = and i64 [[TYPE]], -4 239 // CK0: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[DTYPE]]) 240 // CK0: br label %[[DONE]] 241 // CK0: [[DONE]] 242 // CK0: ret void 243 244 245 // CK0-LABEL: define {{.*}}void @{{.*}}foo{{.*}} 246 void foo(int a){ 247 int i = a; 248 C c; 249 c.a = a; 250 251 // CK0-DAG: call i32 @__tgt_target_mapper(i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZES]]{{.+}}, {{.+}}[[TYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) 252 // CK0-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 253 // CK0-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 254 // CK0-DAG: [[MPRGEP]] = bitcast [1 x i8*]* [[MPR:%[^,]+]] to i8** 255 // CK0-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 256 // CK0-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 257 // CK0-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i[[sz]] 0, i[[sz]] 0 258 // CK0-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.C** 259 // CK0-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** 260 // CK0-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[CBP1]] 261 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[CP1]] 262 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR1]] 263 // CK0: call void [[KERNEL_1:@.+]](%class.C* [[VAL]]) 264 #pragma omp target map(mapper(id),tofrom: c) 265 { 266 ++c.a; 267 } 268 269 // CK0: [[BP2GEP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP2:%[^,]+]], i32 0, i32 0 270 // CK0: [[BP2CAST:%.+]] = bitcast i8** [[BP2GEP]] to %class.C** 271 // CK0: store %class.C* [[CADDR:%[^,]+]], %class.C** [[BP2CAST]], align 272 // CK0: [[P2GEP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_P2:%[^,]+]], i32 0, i32 0 273 // CK0: [[P2CAST:%.+]] = bitcast i8** [[P2GEP]] to %class.C** 274 // CK0: store %class.C* [[CADDR]], %class.C** [[P2CAST]], align 275 // CK0: [[MAPPER2GEP:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_MAPPER2:%[^,]+]], i[[SZ]] 0, i[[SZ]] 0 276 // CK0: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MAPPER2GEP]], align 277 // CK0: [[BP2:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_BP2]], i32 0, i32 0 278 // CK0: [[P2:%.+]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[OFFLOAD_P2]], i32 0, i32 0 279 // CK0: [[MAPPER2:%.+]] = bitcast [1 x i8*]* [[OFFLOAD_MAPPER2]] to i8** 280 // CK0-32: [[TASK:%.+]] = call i8* @__kmpc_omp_target_task_alloc([[IDENT_T]]* {{@.+}}, i32 {{%.+}}, i32 1, i32 40, i32 4, i32 (i32, i8*)* bitcast (i32 (i32, [[KMP_TASK_T_WITH_PRIVATES]]*)* [[TASK_ENTRY:@.+]] to i32 (i32, i8*)*), i64 -1) 281 // CK0-64: [[TASK:%.+]] = call i8* @__kmpc_omp_target_task_alloc([[IDENT_T]]* {{@.+}}, i32 {{%.+}}, i32 1, i64 72, i64 8, i32 (i32, i8*)* bitcast (i32 (i32, [[KMP_TASK_T_WITH_PRIVATES]]*)* [[TASK_ENTRY:@.+]] to i32 (i32, i8*)*), i64 -1) 282 // CK0: [[TASK_CAST:%.+]] = bitcast i8* [[TASK]] to [[KMP_TASK_T_WITH_PRIVATES]]* 283 // CK0: [[TASK_WITH_PRIVATES:%.+]] = getelementptr inbounds [[KMP_TASK_T_WITH_PRIVATES]], [[KMP_TASK_T_WITH_PRIVATES]]* [[TASK_CAST]], i32 0, i32 1 284 // CK0: {{.+}} = call i32 @__kmpc_omp_task([[IDENT_T]]* @1, i32 {{.+}}, i8* [[TASK]]) 285 #pragma omp target map(mapper(id),tofrom: c) nowait 286 { 287 ++c.a; 288 } 289 290 // CK0-DAG: call i32 @__tgt_target_teams_mapper(i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[TEAMSIZES]]{{.+}}, {{.+}}[[TEAMTYPES]]{{.+}}, i8** [[MPRGEP:%.+]], i32 0, i32 0) 291 // CK0-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 292 // CK0-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 293 // CK0-DAG: [[MPRGEP]] = bitcast [1 x i8*]* [[MPR:%[^,]+]] to i8** 294 // CK0-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 295 // CK0-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 296 // CK0-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i[[sz]] 0, i[[sz]] 0 297 // CK0-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.C** 298 // CK0-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** 299 // CK0-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[CBP1]] 300 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[CP1]] 301 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR1]] 302 // CK0: call void [[KERNEL_3:@.+]](%class.C* [[VAL]]) 303 #pragma omp target teams map(mapper(id),to: c) 304 { 305 ++c.a; 306 } 307 308 // CK0-32: [[TASK_1:%.+]] = call i8* @__kmpc_omp_target_task_alloc([[IDENT_T]]* {{@.+}}, i32 {{%.+}}, i32 1, i32 40, i32 4, i32 (i32, i8*)* bitcast (i32 (i32, [[KMP_TASK_T_WITH_PRIVATES_1]]*)* [[TASK_ENTRY_1:@.+]] to i32 (i32, i8*)*), i64 -1) 309 // CK0-64: [[TASK_1:%.+]] = call i8* @__kmpc_omp_target_task_alloc([[IDENT_T]]* {{@.+}}, i32 {{%.+}}, i32 1, i64 72, i64 8, i32 (i32, i8*)* bitcast (i32 (i32, [[KMP_TASK_T_WITH_PRIVATES_1]]*)* [[TASK_ENTRY_1:@.+]] to i32 (i32, i8*)*), i64 -1) 310 // CK0: [[TASK_CAST_1:%.+]] = bitcast i8* [[TASK_1]] to [[KMP_TASK_T_WITH_PRIVATES_1]]* 311 // CK0: [[TASK_CAST_GET_1:%.+]] = getelementptr inbounds [[KMP_TASK_T_WITH_PRIVATES_1]], [[KMP_TASK_T_WITH_PRIVATES_1]]* [[TASK_CAST_1]], i32 0, i32 0 312 // CK0: {{.+}} = getelementptr inbounds [[KMP_TASK_T]], [[KMP_TASK_T]]* [[TASK_CAST_GET_1]], i32 0, i32 0 313 // CK0: {{.+}} = call i32 @__kmpc_omp_task([[IDENT_T]]* @1, i32 {{.+}}, i8* [[TASK_1]]) 314 #pragma omp target teams map(mapper(id),to: c) nowait 315 { 316 ++c.a; 317 } 318 319 // CK0-DAG: call void @__tgt_target_data_begin_mapper(i64 {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[EDSIZES]]{{.+}}, {{.+}}[[EDTYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) 320 // CK0-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 321 // CK0-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 322 // CK0-DAG: [[MPRGEP]] = bitcast [1 x i8*]* [[MPR:%[^,]+]] to i8** 323 // CK0-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 324 // CK0-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 325 // CK0-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i[[sz]] 0, i[[sz]] 0 326 // CK0-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.C** 327 // CK0-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** 328 // CK0-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[CBP1]] 329 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[CP1]] 330 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR1]] 331 #pragma omp target enter data map(mapper(id),to: c) 332 333 // CK0-DAG: call void @__tgt_target_data_begin_nowait_mapper(i64 {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[EDNWSIZES]]{{.+}}, {{.+}}[[EDNWTYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) 334 // CK0-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 335 // CK0-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 336 // CK0-DAG: [[MPRGEP]] = bitcast [1 x i8*]* [[MPR:%[^,]+]] to i8** 337 // CK0-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 338 // CK0-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 339 // CK0-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i[[sz]] 0, i[[sz]] 0 340 // CK0-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.C** 341 // CK0-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** 342 // CK0-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[CBP1]] 343 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[CP1]] 344 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR1]] 345 #pragma omp target enter data map(mapper(id),to: c) nowait 346 347 // CK0-DAG: call void @__tgt_target_data_end_mapper(i64 {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[EXDSIZES]]{{.+}}, {{.+}}[[EXDTYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) 348 // CK0-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 349 // CK0-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 350 // CK0-DAG: [[MPRGEP]] = bitcast [1 x i8*]* [[MPR:%[^,]+]] to i8** 351 // CK0-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 352 // CK0-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 353 // CK0-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i[[sz]] 0, i[[sz]] 0 354 // CK0-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.C** 355 // CK0-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** 356 // CK0-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[CBP1]] 357 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[CP1]] 358 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR1]] 359 #pragma omp target exit data map(mapper(id),from: c) 360 361 // CK0-DAG: call void @__tgt_target_data_end_nowait_mapper(i64 {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[EXDNWSIZES]]{{.+}}, {{.+}}[[EXDNWTYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) 362 // CK0-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 363 // CK0-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 364 // CK0-DAG: [[MPRGEP]] = bitcast [1 x i8*]* [[MPR:%[^,]+]] to i8** 365 // CK0-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 366 // CK0-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 367 // CK0-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i[[sz]] 0, i[[sz]] 0 368 // CK0-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.C** 369 // CK0-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** 370 // CK0-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[CBP1]] 371 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[CP1]] 372 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR1]] 373 #pragma omp target exit data map(mapper(id),from: c) nowait 374 375 // CK0-DAG: call void @__tgt_target_data_update_mapper(i64 -1, i32 1, i8** [[TGEPBP:%.+]], i8** [[TGEPP:%.+]], i64* getelementptr {{.+}}[1 x i64]* [[TSIZES]], i32 0, i32 0), {{.+}}getelementptr {{.+}}[1 x i64]* [[TTYPES]]{{.+}}, i8** [[TMPRGEP:%.+]]) 376 // CK0-DAG: [[TGEPBP]] = getelementptr inbounds {{.+}}[[TBP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 377 // CK0-DAG: [[TGEPP]] = getelementptr inbounds {{.+}}[[TP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 378 // CK0-DAG: [[TMPRGEP]] = bitcast [1 x i8*]* [[TMPR:%[^,]+]] to i8** 379 // CK0-DAG: [[TBP0:%.+]] = getelementptr inbounds {{.+}}[[TBP]], i{{.+}} 0, i{{.+}} 0 380 // CK0-DAG: [[TP0:%.+]] = getelementptr inbounds {{.+}}[[TP]], i{{.+}} 0, i{{.+}} 0 381 // CK0-DAG: [[TMPR1:%.+]] = getelementptr inbounds {{.+}}[[TMPR]], i[[sz]] 0, i[[sz]] 0 382 // CK0-DAG: [[TCBP0:%.+]] = bitcast i8** [[TBP0]] to %class.C** 383 // CK0-DAG: [[TCP0:%.+]] = bitcast i8** [[TP0]] to %class.C** 384 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[TCBP0]] 385 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[TCP0]] 386 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[TMPR1]] 387 #pragma omp target update to(mapper(id): c) 388 389 // CK0-DAG: call void @__tgt_target_data_update_mapper(i64 -1, i32 1, i8** [[FGEPBP:%.+]], i8** [[FGEPP:%.+]], i64* getelementptr {{.+}}[1 x i64]* [[FSIZES]], i32 0, i32 0), {{.+}}getelementptr {{.+}}[1 x i64]* [[FTYPES]]{{.+}}, i8** [[FMPRGEP:%.+]]) 390 // CK0-DAG: [[FGEPBP]] = getelementptr inbounds {{.+}}[[FBP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 391 // CK0-DAG: [[FGEPP]] = getelementptr inbounds {{.+}}[[FP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 392 // CK0-DAG: [[FMPRGEP]] = bitcast [1 x i8*]* [[FMPR:%[^,]+]] to i8** 393 // CK0-DAG: [[FBP0:%.+]] = getelementptr inbounds {{.+}}[[FBP]], i{{.+}} 0, i{{.+}} 0 394 // CK0-DAG: [[FP0:%.+]] = getelementptr inbounds {{.+}}[[FP]], i{{.+}} 0, i{{.+}} 0 395 // CK0-DAG: [[FMPR1:%.+]] = getelementptr inbounds {{.+}}[[FMPR]], i[[sz]] 0, i[[sz]] 0 396 // CK0-DAG: [[FCBP0:%.+]] = bitcast i8** [[FBP0]] to %class.C** 397 // CK0-DAG: [[FCP0:%.+]] = bitcast i8** [[FP0]] to %class.C** 398 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[FCBP0]] 399 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[FCP0]] 400 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[FMPR1]] 401 #pragma omp target update from(mapper(id): c) 402 403 // CK0-DAG: call void @__tgt_target_data_update_nowait_mapper(i64 -1, i32 1, i8** [[FGEPBP:%.+]], i8** [[FGEPP:%.+]], i64* getelementptr {{.+}}[1 x i64]* [[FNWSIZES]], i32 0, i32 0), {{.+}}getelementptr {{.+}}[1 x i64]* [[FNWTYPES]]{{.+}}, i8** [[FMPRGEP:%.+]]) 404 // CK0-DAG: [[FGEPBP]] = getelementptr inbounds {{.+}}[[FBP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 405 // CK0-DAG: [[FGEPP]] = getelementptr inbounds {{.+}}[[FP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 406 // CK0-DAG: [[FMPRGEP]] = bitcast [1 x i8*]* [[FMPR:%[^,]+]] to i8** 407 // CK0-DAG: [[FBP0:%.+]] = getelementptr inbounds {{.+}}[[FBP]], i{{.+}} 0, i{{.+}} 0 408 // CK0-DAG: [[FP0:%.+]] = getelementptr inbounds {{.+}}[[FP]], i{{.+}} 0, i{{.+}} 0 409 // CK0-DAG: [[FMPR1:%.+]] = getelementptr inbounds {{.+}}[[FMPR]], i[[sz]] 0, i[[sz]] 0 410 // CK0-DAG: [[FCBP0:%.+]] = bitcast i8** [[FBP0]] to %class.C** 411 // CK0-DAG: [[FCP0:%.+]] = bitcast i8** [[FP0]] to %class.C** 412 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[FCBP0]] 413 // CK0-DAG: store %class.C* [[VAL]], %class.C** [[FCP0]] 414 // CK0-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[FMPR1]] 415 #pragma omp target update from(mapper(id): c) nowait 416 } 417 418 419 // CK0: define internal void [[KERNEL_1]](%class.C* {{.+}}[[ARG:%.+]]) 420 // CK0: [[ADDR:%.+]] = alloca %class.C*, 421 // CK0: store %class.C* [[ARG]], %class.C** [[ADDR]] 422 // CK0: [[CADDR:%.+]] = load %class.C*, %class.C** [[ADDR]] 423 // CK0: [[CAADDR:%.+]] = getelementptr inbounds %class.C, %class.C* [[CADDR]], i32 0, i32 0 424 // CK0: [[VAL:%[^,]+]] = load i32, i32* [[CAADDR]] 425 // CK0: {{.+}} = add nsw i32 [[VAL]], 1 426 // CK0: } 427 428 // CK0: define internal void [[KERNEL_2:@.+]](%class.C* {{.+}}[[ARG:%.+]]) 429 // CK0: [[ADDR:%.+]] = alloca %class.C*, 430 // CK0: store %class.C* [[ARG]], %class.C** [[ADDR]] 431 // CK0: [[CADDR:%.+]] = load %class.C*, %class.C** [[ADDR]] 432 // CK0: [[CAADDR:%.+]] = getelementptr inbounds %class.C, %class.C* [[CADDR]], i32 0, i32 0 433 // CK0: [[VAL:%[^,]+]] = load i32, i32* [[CAADDR]] 434 // CK0: {{.+}} = add nsw i32 [[VAL]], 1 435 // CK0: } 436 437 // CK0: define internal void [[OUTLINED:@.+]](i32 {{.*}}{{[^,]+}}, [[ANON_T]]* noalias [[CTXARG:%.+]]) 438 // CK0-DAG: call i32 @__tgt_target_nowait_mapper(i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], {{.+}}[[SIZEGEP:%[0-9]+]], {{.+}}[[NWTYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) 439 // CK0-DAG: [[BPGEP]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[BPFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 440 // CK0-DAG: [[PGEP]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[PFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 441 // CK0-DAG: [[SIZEGEP]] = getelementptr inbounds [1 x i64], [1 x i64]* [[SIZEFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 442 // CK0-DAG: [[MPRGEP]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[MPRFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 443 // CK0-DAG: [[BPFPADDR]] = load [1 x i8*]*, [1 x i8*]** [[FPPTRADDR_BP:%.+]], align 444 // CK0-DAG: [[PFPADDR]] = load [1 x i8*]*, [1 x i8*]** [[FPPTRADDR_P:%.+]], align 445 // CK0-DAG: [[SIZEFPADDR]] = load [1 x i64]*, [1 x i64]** [[FPPTRADDR_SIZE:%.+]], align 446 // CK0-DAG: [[MPRFPADDR]] = load [1 x i8*]*, [1 x i8*]** [[FPPTRADDR_MPR:%.+]], align 447 // CK0-DAG: call void (i8*, ...) %1(i8* %2, {{.+}}[[FPPTRADDR_BP]], {{.+}}[[FPPTRADDR_P]], {{.+}}[[FPPTRADDR_SIZE]], {{.+}}[[FPPTRADDR_MPR]]) 448 // CK0-DAG: call void [[KERNEL_2:@.+]](%class.C* [[KERNELARG:%.+]]) 449 // CK0-DAG: [[KERNELARG]] = load %class.C*, %class.C** [[KERNELARGGEP:%.+]], align 450 // CK0-DAG: [[KERNELARGGEP]] = getelementptr inbounds [[ANON_T]], [[ANON_T]]* [[CTX:%.+]], i32 0, i32 0 451 // CK0-DAG: [[CTX]] = load [[ANON_T]]*, [[ANON_T]]** [[CTXADDR:%.+]], align 452 // CK0-DAG: store [[ANON_T]]* [[CTXARG]], [[ANON_T]]** [[CTXADDR]], align 453 // CK0: } 454 455 // CK0: define internal {{.*}}i32 [[TASK_ENTRY]](i32 {{.*}}%0, [[KMP_TASK_T_WITH_PRIVATES]]* noalias %1) 456 // CK0: store [[KMP_TASK_T_WITH_PRIVATES]]* %1, [[KMP_TASK_T_WITH_PRIVATES]]** [[ADDR:%.+]], align 457 // CK0: [[TASK_T_WITH_PRIVATES:%.+]] = load [[KMP_TASK_T_WITH_PRIVATES]]*, [[KMP_TASK_T_WITH_PRIVATES]]** [[ADDR]], align 458 // CK0: [[TASKGEP:%.+]] = getelementptr inbounds [[KMP_TASK_T_WITH_PRIVATES]], [[KMP_TASK_T_WITH_PRIVATES]]* [[TASK_T_WITH_PRIVATES]], i32 0, i32 0 459 // CK0: [[SHAREDSGEP:%.+]] = getelementptr inbounds [[KMP_TASK_T]], [[KMP_TASK_T]]* [[TASKGEP]], i32 0, i32 0 460 // CK0: [[SHAREDS:%.+]] = load i8*, i8** [[SHAREDSGEP]], align 461 // CK0: [[ANON:%.+]] = bitcast i8* [[SHAREDS]] to [[ANON_T]]* 462 // CK0: [[PRIVATESGEP:%.+]] = getelementptr inbounds [[KMP_TASK_T_WITH_PRIVATES]], [[KMP_TASK_T_WITH_PRIVATES]]* [[TASK_T_WITH_PRIVATES]], i32 0, i32 1 463 // CK0: [[PRIVATES:%.+]] = bitcast [[KMP_PRIVATES_T]]* [[PRIVATESGEP]] to i8* 464 // CK0: [[TASK_WITH_PRIVATES:%.+]] = bitcast [[KMP_TASK_T_WITH_PRIVATES]]* [[TASK_T_WITH_PRIVATES]] to i8* 465 // CK0: call void [[OUTLINED]](i32 {{%.+}}, i32* {{%.+}}, i8* [[PRIVATES]], {{.+}}, i8* [[TASK_WITH_PRIVATES]], [[ANON_T]]* [[ANON]]) 466 // CK0: } 467 468 // CK0: define internal void [[OUTLINE_1:@.+]](i32 {{.*}}%.global_tid.{{.+}}, [[ANON_T_0]]* noalias [[CTXARG:%.+]]) 469 // CK0-DAG: call i32 @__tgt_target_teams_nowait_mapper(i64 {{.+}}, i8* {{.+}}, i32 1, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], i64* [[SIZEGEP:%[0-9]+]], {{.+}}[[TEAMNWTYPES]]{{.+}}, i8** [[MPRGEP:%.+]], i32 0, i32 0) 470 // CK0-DAG: [[BPGEP]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[BPFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 471 // CK0-DAG: [[PGEP]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[PFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 472 // CK0-DAG: [[SIZEGEP]] = getelementptr inbounds [1 x i64], [1 x i64]* [[SIZEFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 473 // CK0-DAG: [[MPRGEP]] = getelementptr inbounds [1 x i8*], [1 x i8*]* [[MPRFPADDR:%.+]], i[[SZ]] 0, i[[SZ]] 0 474 // CK0-DAG: [[BPFPADDR]] = load [1 x i8*]*, [1 x i8*]** [[FPPTRADDR_BP:%.+]], align 475 // CK0-DAG: [[PFPADDR]] = load [1 x i8*]*, [1 x i8*]** [[FPPTRADDR_P:%.+]], align 476 // CK0-DAG: [[SIZEFPADDR]] = load [1 x i64]*, [1 x i64]** [[FPPTRADDR_SIZE:%.+]], align 477 // CK0-DAG: [[MPRFPADDR]] = load [1 x i8*]*, [1 x i8*]** [[FPPTRADDR_MPR:%.+]], align 478 // CK0-DAG: call void (i8*, ...) %1(i8* %2, {{.+}}[[FPPTRADDR_BP]], {{.+}}[[FPPTRADDR_P]], {{.+}}[[FPPTRADDR_SIZE]], {{.+}}[[FPPTRADDR_MPR]]) 479 // CK0-DAG: call void [[KERNEL_2:@.+]](%class.C* [[KERNELARG:%.+]]) 480 // CK0-DAG: [[KERNELARG]] = load %class.C*, %class.C** [[KERNELARGGEP:%.+]], align 481 // CK0-DAG: [[KERNELARGGEP]] = getelementptr inbounds [[ANON_T_0]], [[ANON_T_0]]* [[CTX:%.+]], i32 0, i32 0 482 // CK0-DAG: [[CTX]] = load [[ANON_T_0]]*, [[ANON_T_0]]** [[CTXADDR:%.+]], align 483 // CK0-DAG: store [[ANON_T_0]]* [[CTXARG]], [[ANON_T_0]]** [[CTXADDR]], align 484 // CK0: } 485 486 // CK0: define internal {{.*}}i32 [[TASK_ENTRY_1]](i32 {{.*}}%0, [[KMP_TASK_T_WITH_PRIVATES_1]]* noalias %1) 487 // CK0: store [[KMP_TASK_T_WITH_PRIVATES_1]]* %1, [[KMP_TASK_T_WITH_PRIVATES_1]]** [[ADDR:%.+]], align 488 // CK0: [[TASK_T_WITH_PRIVATES:%.+]] = load [[KMP_TASK_T_WITH_PRIVATES_1]]*, [[KMP_TASK_T_WITH_PRIVATES_1]]** [[ADDR]], align 489 // CK0: [[TASKGEP:%.+]] = getelementptr inbounds [[KMP_TASK_T_WITH_PRIVATES_1]], [[KMP_TASK_T_WITH_PRIVATES_1]]* [[TASK_T_WITH_PRIVATES]], i32 0, i32 0 490 // CK0: [[SHAREDSGEP:%.+]] = getelementptr inbounds [[KMP_TASK_T]], [[KMP_TASK_T]]* [[TASKGEP]], i32 0, i32 0 491 // CK0: [[SHAREDS:%.+]] = load i8*, i8** [[SHAREDSGEP]], align 492 // CK0: [[ANON:%.+]] = bitcast i8* [[SHAREDS]] to [[ANON_T_0]]* 493 // CK0: [[PRIVATESGEP:%.+]] = getelementptr inbounds [[KMP_TASK_T_WITH_PRIVATES_1]], [[KMP_TASK_T_WITH_PRIVATES_1]]* [[TASK_T_WITH_PRIVATES]], i32 0, i32 1 494 // CK0: [[PRIVATES:%.+]] = bitcast [[KMP_PRIVATES_T_2]]* [[PRIVATESGEP]] to i8* 495 // CK0: [[TASK_WITH_PRIVATES:%.+]] = bitcast [[KMP_TASK_T_WITH_PRIVATES_1]]* [[TASK_T_WITH_PRIVATES]] to i8* 496 // CK0: call void [[OUTLINE_1]](i32 {{%.+}}, i32* {{%.+}}, i8* [[PRIVATES]], {{.+}}, i8* [[TASK_WITH_PRIVATES]], [[ANON_T_0]]* [[ANON]]) 497 // CK0: } 498 499 #endif // CK0 500 501 502 ///==========================================================================/// 503 // RUN: %clang_cc1 -DCK1 -verify -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK1 --check-prefix CK1-64 %s 504 // RUN: %clang_cc1 -DCK1 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 505 // RUN: %clang_cc1 -DCK1 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK1 --check-prefix CK1-64 %s 506 // RUN: %clang_cc1 -DCK1 -verify -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK1 --check-prefix CK1-32 %s 507 // RUN: %clang_cc1 -DCK1 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 508 // RUN: %clang_cc1 -DCK1 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK1 --check-prefix CK1-32 %s 509 510 // RUN: %clang_cc1 -DCK1 -verify -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 511 // RUN: %clang_cc1 -DCK1 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 512 // RUN: %clang_cc1 -DCK1 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 513 // RUN: %clang_cc1 -DCK1 -verify -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 514 // RUN: %clang_cc1 -DCK1 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 515 // RUN: %clang_cc1 -DCK1 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 516 517 #ifdef CK1 518 // C++ template 519 520 template <class T> 521 class C { 522 public: 523 T a; 524 }; 525 526 #pragma omp declare mapper(id: C<int> s) map(s.a) 527 528 // CK1-LABEL: define {{.*}}void @.omp_mapper.{{.*}}C{{.*}}.id{{.*}}(i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}}) 529 // CK1: store i8* %{{[^,]+}}, i8** [[HANDLEADDR:%[^,]+]] 530 // CK1: store i8* %{{[^,]+}}, i8** [[BPTRADDR:%[^,]+]] 531 // CK1: store i8* %{{[^,]+}}, i8** [[VPTRADDR:%[^,]+]] 532 // CK1: store i64 %{{[^,]+}}, i{{64|32}}* [[SIZEADDR:%[^,]+]] 533 // CK1: store i64 %{{[^,]+}}, i64* [[TYPEADDR:%[^,]+]] 534 // CK1-DAG: [[BYTESIZE:%.+]] = load i64, i64* [[SIZEADDR]] 535 // CK1-DAG: [[SIZE:%.+]] = udiv exact i64 [[BYTESIZE]], 4 536 // CK1-DAG: [[TYPE:%.+]] = load i64, i64* [[TYPEADDR]] 537 // CK1-DAG: [[HANDLE:%.+]] = load i8*, i8** [[HANDLEADDR]] 538 // CK1-DAG: [[PTRBEGIN:%.+]] = bitcast i8** [[VPTRADDR]] to %class.C** 539 // CK1-DAG: [[PTREND:%.+]] = getelementptr %class.C*, %class.C** [[PTRBEGIN]], i64 [[SIZE]] 540 // CK1-DAG: [[BPTR:%.+]] = load i8*, i8** [[BPTRADDR]] 541 // CK1-DAG: [[BEGIN:%.+]] = load i8*, i8** [[VPTRADDR]] 542 // CK1: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 543 // CK1: br i1 [[ISARRAY]], label %[[INITEVALDEL:[^,]+]], label %[[LHEAD:[^,]+]] 544 545 // CK1: [[INITEVALDEL]] 546 // CK1: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 547 // CK1: [[ISNOTDEL:%.+]] = icmp eq i64 [[TYPEDEL]], 0 548 // CK1: br i1 [[ISNOTDEL]], label %[[INIT:[^,]+]], label %[[LHEAD:[^,]+]] 549 // CK1: [[INIT]] 550 // CK1-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 4 551 // CK1-DAG: [[ITYPE:%.+]] = and i64 [[TYPE]], -4 552 // CK1: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[ITYPE]]) 553 // CK1: br label %[[LHEAD:[^,]+]] 554 555 // CK1: [[LHEAD]] 556 // CK1: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]] 557 // CK1: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]] 558 // CK1: [[LBODY]] 559 // CK1: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ] 560 // CK1: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]] 561 // CK1-DAG: [[ABEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 0 562 // CK1-DAG: [[AEND:%.+]] = getelementptr i32, i32* [[ABEGIN]], i32 1 563 // CK1-DAG: [[ABEGINV:%.+]] = bitcast i32* [[ABEGIN]] to i8* 564 // CK1-DAG: [[AENDV:%.+]] = bitcast i32* [[AEND]] to i8* 565 // CK1-DAG: [[ABEGINI:%.+]] = ptrtoint i8* [[ABEGINV]] to i64 566 // CK1-DAG: [[AENDI:%.+]] = ptrtoint i8* [[AENDV]] to i64 567 // CK1-DAG: [[CSIZE:%.+]] = sub i64 [[AENDI]], [[ABEGINI]] 568 // CK1-DAG: [[CUSIZE:%.+]] = sdiv exact i64 [[CSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64) 569 // CK1-DAG: [[BPTRADDR0BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 570 // CK1-DAG: [[PTRADDR0BC:%.+]] = bitcast i32* [[ABEGIN]] to i8* 571 // CK1-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]]) 572 // CK1-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48 573 // CK1-DAG: br label %[[MEMBER:[^,]+]] 574 // CK1-DAG: [[MEMBER]] 575 // CK1-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 576 // CK1-DAG: [[MEMBERCOM]] 577 // CK1-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]] 578 // CK1-DAG: br label %[[LTYPE]] 579 // CK1-DAG: [[LTYPE]] 580 // CK1-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 581 // CK1-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 582 // CK1-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 583 // CK1-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 584 // CK1-DAG: [[ALLOC]] 585 // CK1-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 586 // CK1-DAG: br label %[[TYEND:[^,]+]] 587 // CK1-DAG: [[ALLOCELSE]] 588 // CK1-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 589 // CK1-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 590 // CK1-DAG: [[TO]] 591 // CK1-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 592 // CK1-DAG: br label %[[TYEND]] 593 // CK1-DAG: [[TOELSE]] 594 // CK1-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 595 // CK1-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 596 // CK1-DAG: [[FROM]] 597 // CK1-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 598 // CK1-DAG: br label %[[TYEND]] 599 // CK1-DAG: [[TYEND]] 600 // CK1-DAG: [[TYPE0:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 601 // CK1-64: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR0BC]], i8* [[PTRADDR0BC]], i64 [[CUSIZE]], i64 [[TYPE0]]) 602 // CK1-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 603 // CK1-DAG: [[PTRADDR1BC:%.+]] = bitcast i32* [[ABEGIN]] to i8* 604 // CK1-DAG: br label %[[MEMBER:[^,]+]] 605 // CK1-DAG: [[MEMBER]] 606 // CK1-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 607 // CK1-DAG: [[MEMBERCOM]] 608 // 281474976710659 == 0x1,000,000,003 609 // CK1-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710659, [[SHIPRESIZE]] 610 // CK1-DAG: br label %[[LTYPE]] 611 // CK1-DAG: [[LTYPE]] 612 // CK1-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710659, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 613 // CK1-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 614 // CK1-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 615 // CK1-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 616 // CK1-DAG: [[ALLOC]] 617 // CK1-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 618 // CK1-DAG: br label %[[TYEND:[^,]+]] 619 // CK1-DAG: [[ALLOCELSE]] 620 // CK1-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 621 // CK1-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 622 // CK1-DAG: [[TO]] 623 // CK1-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 624 // CK1-DAG: br label %[[TYEND]] 625 // CK1-DAG: [[TOELSE]] 626 // CK1-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 627 // CK1-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 628 // CK1-DAG: [[FROM]] 629 // CK1-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 630 // CK1-DAG: br label %[[TYEND]] 631 // CK1-DAG: [[TYEND]] 632 // CK1-DAG: [[TYPE1:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 633 // CK1: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR1BC]], i8* [[PTRADDR1BC]], i64 4, i64 [[TYPE1]]) 634 // CK1: [[PTRNEXT]] = getelementptr %class.C*, %class.C** [[PTR]], i32 1 635 // CK1: [[ISDONE:%.+]] = icmp eq %class.C** [[PTRNEXT]], [[PTREND]] 636 // CK1: br i1 [[ISDONE]], label %[[LEXIT:[^,]+]], label %[[LBODY]] 637 638 // CK1: [[LEXIT]] 639 // CK1: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 640 // CK1: br i1 [[ISARRAY]], label %[[EVALDEL:[^,]+]], label %[[DONE]] 641 // CK1: [[EVALDEL]] 642 // CK1: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 643 // CK1: [[ISDEL:%.+]] = icmp ne i64 [[TYPEDEL]], 0 644 // CK1: br i1 [[ISDEL]], label %[[DEL:[^,]+]], label %[[DONE]] 645 // CK1: [[DEL]] 646 // CK1-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 4 647 // CK1-DAG: [[DTYPE:%.+]] = and i64 [[TYPE]], -4 648 // CK1: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[DTYPE]]) 649 // CK1: br label %[[DONE]] 650 // CK1: [[DONE]] 651 // CK1: ret void 652 653 #endif // CK1 654 655 656 ///==========================================================================/// 657 // RUN: %clang_cc1 -DCK2 -verify -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK2 --check-prefix CK2-64 %s 658 // RUN: %clang_cc1 -DCK2 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 659 // RUN: %clang_cc1 -DCK2 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK2 --check-prefix CK2-64 %s 660 // RUN: %clang_cc1 -DCK2 -verify -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK2 --check-prefix CK2-32 %s 661 // RUN: %clang_cc1 -DCK2 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 662 // RUN: %clang_cc1 -DCK2 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK2 --check-prefix CK2-32 %s 663 664 // RUN: %clang_cc1 -DCK2 -verify -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 665 // RUN: %clang_cc1 -DCK2 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 666 // RUN: %clang_cc1 -DCK2 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 667 // RUN: %clang_cc1 -DCK2 -verify -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 668 // RUN: %clang_cc1 -DCK2 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 669 // RUN: %clang_cc1 -DCK2 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 670 671 #ifdef CK2 672 // Nested mappers. 673 674 class B { 675 public: 676 double a; 677 }; 678 679 class C { 680 public: 681 double a; 682 B b; 683 }; 684 685 #pragma omp declare mapper(B s) map(s.a) 686 687 #pragma omp declare mapper(id: C s) map(s.b) 688 689 // CK2: define {{.*}}void [[BMPRFUNC:@[.]omp_mapper[.].*B[.]default]](i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}}) 690 691 // CK2-LABEL: define {{.*}}void @.omp_mapper.{{.*}}C{{.*}}.id(i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}}) 692 // CK2: store i8* %{{[^,]+}}, i8** [[HANDLEADDR:%[^,]+]] 693 // CK2: store i8* %{{[^,]+}}, i8** [[BPTRADDR:%[^,]+]] 694 // CK2: store i8* %{{[^,]+}}, i8** [[VPTRADDR:%[^,]+]] 695 // CK2: store i64 %{{[^,]+}}, i{{64|32}}* [[SIZEADDR:%[^,]+]] 696 // CK2: store i64 %{{[^,]+}}, i64* [[TYPEADDR:%[^,]+]] 697 // CK2-DAG: [[BYTESIZE:%.+]] = load i64, i64* [[SIZEADDR]] 698 // CK2-DAG: [[SIZE:%.+]] = udiv exact i64 [[BYTESIZE]], 16 699 // CK2-DAG: [[TYPE:%.+]] = load i64, i64* [[TYPEADDR]] 700 // CK2-DAG: [[HANDLE:%.+]] = load i8*, i8** [[HANDLEADDR]] 701 // CK2-DAG: [[PTRBEGIN:%.+]] = bitcast i8** [[VPTRADDR]] to %class.C** 702 // CK2-DAG: [[PTREND:%.+]] = getelementptr %class.C*, %class.C** [[PTRBEGIN]], i64 [[SIZE]] 703 // CK2-DAG: [[BPTR:%.+]] = load i8*, i8** [[BPTRADDR]] 704 // CK2-DAG: [[BEGIN:%.+]] = load i8*, i8** [[VPTRADDR]] 705 // CK2: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 706 // CK2: br i1 [[ISARRAY]], label %[[INITEVALDEL:[^,]+]], label %[[LHEAD:[^,]+]] 707 708 // CK2: [[INITEVALDEL]] 709 // CK2: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 710 // CK2: [[ISNOTDEL:%.+]] = icmp eq i64 [[TYPEDEL]], 0 711 // CK2: br i1 [[ISNOTDEL]], label %[[INIT:[^,]+]], label %[[LHEAD:[^,]+]] 712 // CK2: [[INIT]] 713 // CK2-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 16 714 // CK2-DAG: [[ITYPE:%.+]] = and i64 [[TYPE]], -4 715 // CK2: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[ITYPE]]) 716 // CK2: br label %[[LHEAD:[^,]+]] 717 718 // CK2: [[LHEAD]] 719 // CK2: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]] 720 // CK2: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]] 721 // CK2: [[LBODY]] 722 // CK2: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ] 723 // CK2: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]] 724 // CK2-DAG: [[BBEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 1 725 // CK2-DAG: [[BEND:%.+]] = getelementptr %class.B, %class.B* [[BBEGIN]], i32 1 726 // CK2-DAG: [[BBEGINV:%.+]] = bitcast %class.B* [[BBEGIN]] to i8* 727 // CK2-DAG: [[BENDV:%.+]] = bitcast %class.B* [[BEND]] to i8* 728 // CK2-DAG: [[BBEGINI:%.+]] = ptrtoint i8* [[BBEGINV]] to i64 729 // CK2-DAG: [[BENDI:%.+]] = ptrtoint i8* [[BENDV]] to i64 730 // CK2-DAG: [[BSIZE:%.+]] = sub i64 [[BENDI]], [[BBEGINI]] 731 // CK2-DAG: [[BUSIZE:%.+]] = sdiv exact i64 [[BSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64) 732 // CK2-DAG: [[BPTRADDR0BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 733 // CK2-DAG: [[PTRADDR0BC:%.+]] = bitcast %class.B* [[BBEGIN]] to i8* 734 // CK2-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]]) 735 // CK2-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48 736 // CK2-DAG: br label %[[MEMBER:[^,]+]] 737 // CK2-DAG: [[MEMBER]] 738 // CK2-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 739 // CK2-DAG: [[MEMBERCOM]] 740 // CK2-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]] 741 // CK2-DAG: br label %[[LTYPE]] 742 // CK2-DAG: [[LTYPE]] 743 // CK2-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 744 // CK2-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 745 // CK2-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 746 // CK2-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 747 // CK2-DAG: [[ALLOC]] 748 // CK2-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 749 // CK2-DAG: br label %[[TYEND:[^,]+]] 750 // CK2-DAG: [[ALLOCELSE]] 751 // CK2-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 752 // CK2-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 753 // CK2-DAG: [[TO]] 754 // CK2-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 755 // CK2-DAG: br label %[[TYEND]] 756 // CK2-DAG: [[TOELSE]] 757 // CK2-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 758 // CK2-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 759 // CK2-DAG: [[FROM]] 760 // CK2-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 761 // CK2-DAG: br label %[[TYEND]] 762 // CK2-DAG: [[TYEND]] 763 // CK2-DAG: [[TYPE0:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 764 // CK2-64: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR0BC]], i8* [[PTRADDR0BC]], i64 [[BUSIZE]], i64 [[TYPE0]]) 765 // CK2-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 766 // CK2-DAG: [[PTRADDR1BC:%.+]] = bitcast %class.B* [[BBEGIN]] to i8* 767 // CK2-DAG: br label %[[MEMBER:[^,]+]] 768 // CK2-DAG: [[MEMBER]] 769 // CK2-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 770 // CK2-DAG: [[MEMBERCOM]] 771 // 281474976710659 == 0x1,000,000,003 772 // CK2-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710659, [[SHIPRESIZE]] 773 // CK2-DAG: br label %[[LTYPE]] 774 // CK2-DAG: [[LTYPE]] 775 // CK2-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710659, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 776 // CK2-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 777 // CK2-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 778 // CK2-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 779 // CK2-DAG: [[ALLOC]] 780 // CK2-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 781 // CK2-DAG: br label %[[TYEND:[^,]+]] 782 // CK2-DAG: [[ALLOCELSE]] 783 // CK2-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 784 // CK2-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 785 // CK2-DAG: [[TO]] 786 // CK2-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 787 // CK2-DAG: br label %[[TYEND]] 788 // CK2-DAG: [[TOELSE]] 789 // CK2-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 790 // CK2-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 791 // CK2-DAG: [[FROM]] 792 // CK2-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 793 // CK2-DAG: br label %[[TYEND]] 794 // CK2-DAG: [[TYEND]] 795 // CK2-DAG: [[TYPE1:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 796 // CK2: call void [[BMPRFUNC]](i8* [[HANDLE]], i8* [[BPTRADDR1BC]], i8* [[PTRADDR1BC]], i64 8, i64 [[TYPE1]]) 797 // CK2: [[PTRNEXT]] = getelementptr %class.C*, %class.C** [[PTR]], i32 1 798 // CK2: [[ISDONE:%.+]] = icmp eq %class.C** [[PTRNEXT]], [[PTREND]] 799 // CK2: br i1 [[ISDONE]], label %[[LEXIT:[^,]+]], label %[[LBODY]] 800 801 // CK2: [[LEXIT]] 802 // CK2: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 803 // CK2: br i1 [[ISARRAY]], label %[[EVALDEL:[^,]+]], label %[[DONE]] 804 // CK2: [[EVALDEL]] 805 // CK2: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 806 // CK2: [[ISDEL:%.+]] = icmp ne i64 [[TYPEDEL]], 0 807 // CK2: br i1 [[ISDEL]], label %[[DEL:[^,]+]], label %[[DONE]] 808 // CK2: [[DEL]] 809 // CK2-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 16 810 // CK2-DAG: [[DTYPE:%.+]] = and i64 [[TYPE]], -4 811 // CK2: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[DTYPE]]) 812 // CK2: br label %[[DONE]] 813 // CK2: [[DONE]] 814 // CK2: ret void 815 816 #endif // CK2 817 818 819 ///==========================================================================/// 820 // RUN: %clang_cc1 -DCK3 -verify -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK3 %s 821 // RUN: %clang_cc1 -DCK3 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 822 // RUN: %clang_cc1 -DCK3 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK3 %s 823 // RUN: %clang_cc1 -DCK3 -verify -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK3 %s 824 // RUN: %clang_cc1 -DCK3 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 825 // RUN: %clang_cc1 -DCK3 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK3 %s 826 827 // RUN: %clang_cc1 -DCK3 -verify -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 828 // RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 829 // RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 830 // RUN: %clang_cc1 -DCK3 -verify -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 831 // RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 832 // RUN: %clang_cc1 -DCK3 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 833 834 #ifdef CK3 835 // map of array sections and nested components. 836 837 // CK3-LABEL: @.__omp_offloading_{{.*}}foo{{.*}}.region_id = weak constant i8 0 838 // CK3: [[TYPES:@.+]] = {{.+}}constant [3 x i64] [i64 32, i64 281474976710659, i64 35] 839 840 class C { 841 public: 842 int a; 843 double *b; 844 }; 845 846 class B { 847 public: 848 C c; 849 }; 850 851 #pragma omp declare mapper(id: C s) map(s.a, s.b[0:2]) 852 853 // CK3: define {{.*}}void [[MPRFUNC:@[.]omp_mapper[.].*C[.]id]](i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}}) 854 855 // CK3-LABEL: define {{.*}}void @{{.*}}foo{{.*}} 856 void foo(int a){ 857 // CK3-DAG: [[CVAL:%.+]] = alloca [10 x %class.C] 858 // CK3-DAG: [[BVAL:%.+]] = alloca %class.B 859 C c[10]; 860 B b; 861 862 // CK3-DAG: [[BC:%.+]] = getelementptr inbounds %class.B, %class.B* [[BVAL]], i32 0, i32 0 863 // CK3-DAG: [[BCEND:%.+]] = getelementptr %class.C, %class.C* [[BC]], i32 1 864 // CK3-DAG: [[BCC:%.+]] = bitcast %class.C* [[BC]] to i8* 865 // CK3-DAG: [[BCENDC:%.+]] = bitcast %class.C* [[BCEND]] to i8* 866 // CK3-DAG: [[BCI:%.+]] = ptrtoint i8* [[BCC]] to i64 867 // CK3-DAG: [[BCENDI:%.+]] = ptrtoint i8* [[BCENDC]] to i64 868 // CK3-DAG: [[BSIZE:%.+]] = sub i64 [[BCENDI]], [[BCI]] 869 // CK3-DAG: [[BSIZED:%.+]] = sdiv exact i64 [[BSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64) 870 871 // CK3-DAG: call i32 @__tgt_target_mapper(i64 {{.+}}, i8* {{.+}}, i32 3, i8** [[BPGEP:%[0-9]+]], i8** [[PGEP:%[0-9]+]], i64* [[SGEP:%[^,]+]], {{.+}}[[TYPES]]{{.+}}, i8** [[MPRGEP:%.+]]) 872 // CK3-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0 873 // CK3-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0 874 // CK3-DAG: [[SGEP]] = getelementptr inbounds {{.+}}[[SIZES:%[^,]+]], i32 0, i32 0 875 // CK3-DAG: [[MPRGEP]] = bitcast [3 x i8*]* [[MPR:%[^,]+]] to i8** 876 // CK3-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0 877 // CK3-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0 878 // CK3-DAG: [[S1:%.+]] = getelementptr inbounds {{.+}}[[SIZES]], i32 0, i32 0 879 // CK3-DAG: [[MPR1:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 0 880 // CK3-DAG: [[CBP1:%.+]] = bitcast i8** [[BP1]] to %class.B** 881 // CK3-DAG: [[CP1:%.+]] = bitcast i8** [[P1]] to %class.C** 882 // CK3-DAG: store %class.B* [[BVAL]], %class.B** [[CBP1]] 883 // CK3-DAG: store %class.C* [[BC]], %class.C** [[CP1]] 884 // CK3-DAG: store i64 [[BSIZED]], i64* [[S1]] 885 // CK3-DAG: store i8* null, i8** [[MPR1]] 886 // CK3-DAG: [[BP2:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 1 887 // CK3-DAG: [[P2:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 1 888 // CK3-DAG: [[S2:%.+]] = getelementptr inbounds {{.+}}[[SIZES]], i32 0, i32 1 889 // CK3-DAG: [[MPR2:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 1 890 // CK3-DAG: [[CBP2:%.+]] = bitcast i8** [[BP2]] to %class.B** 891 // CK3-DAG: [[CP2:%.+]] = bitcast i8** [[P2]] to %class.C** 892 // CK3-DAG: store %class.B* [[BVAL]], %class.B** [[CBP2]] 893 // CK3-DAG: store %class.C* [[BC]], %class.C** [[CP2]] 894 // CK3-64-DAG: store i64 16, i64* [[S2]] 895 // CK3-32-DAG: store i64 8, i64* [[S2]] 896 // CK3-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR2]] 897 // CK3-DAG: [[BP3:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 2 898 // CK3-DAG: [[P3:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 2 899 // CK3-DAG: [[S3:%.+]] = getelementptr inbounds {{.+}}[[SIZES]], i32 0, i32 2 900 // CK3-DAG: [[MPR3:%.+]] = getelementptr inbounds {{.+}}[[MPR]], i{{64|32}} 0, i{{64|32}} 2 901 // CK3-DAG: [[CBP3:%.+]] = bitcast i8** [[BP3]] to [10 x %class.C]** 902 // CK3-DAG: [[CP3:%.+]] = bitcast i8** [[P3]] to %class.C** 903 // CK3-DAG: store [10 x %class.C]* [[CVAL]], [10 x %class.C]** [[CBP3]] 904 // CK3-DAG: [[CVALGEP:%.+]] = getelementptr inbounds {{.+}}[[CVAL]], i{{64|32}} 0, i{{64|32}} 0 905 // CK3-DAG: store %class.C* [[CVALGEP]], %class.C** [[CP3]] 906 // CK3-64-DAG: store i64 160, i64* [[S3]] 907 // CK3-32-DAG: store i64 80, i64* [[S3]] 908 // CK3-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[MPR3]] 909 // CK3: call void [[KERNEL:@.+]](%class.B* [[BVAL]], [10 x %class.C]* [[CVAL]]) 910 #pragma omp target map(mapper(id),tofrom: c[0:10], b.c) 911 for (int i = 0; i < 10; i++) { 912 b.c.a += ++c[i].a; 913 } 914 } 915 916 917 // CK3: define internal void [[KERNEL]](%class.B* {{[^,]+}}, [10 x %class.C]* {{[^,]+}}) 918 919 #endif // CK3 920 921 ///==========================================================================/// 922 // RUN: %clang_cc1 -DCK4 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK4 --check-prefix CK4-64 %s 923 // RUN: %clang_cc1 -DCK4 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 924 // RUN: %clang_cc1 -DCK4 -fopenmp -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK4 --check-prefix CK4-64 %s 925 // RUN: %clang_cc1 -DCK4 -verify -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix CK4 --check-prefix CK4-32 %s 926 // RUN: %clang_cc1 -DCK4 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 927 // RUN: %clang_cc1 -DCK4 -fopenmp -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix CK4 --check-prefix CK4-32 %s 928 929 // RUN: %clang_cc1 -DCK4 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 930 // RUN: %clang_cc1 -DCK4 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 931 // RUN: %clang_cc1 -DCK4 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -triple powerpc64le-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 932 // RUN: %clang_cc1 -DCK4 -verify -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -emit-llvm -femit-all-decls -disable-llvm-passes %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 933 // RUN: %clang_cc1 -DCK4 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -femit-all-decls -disable-llvm-passes -o %t %s 934 // RUN: %clang_cc1 -DCK4 -fopenmp-simd -fopenmp-version=51 -fopenmp-targets=i386-pc-linux-gnu -x c++ -triple i386-unknown-unknown -std=c++11 -femit-all-decls -disable-llvm-passes -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 935 936 #ifdef CK4 937 // Mapper function code generation and runtime interface. 938 939 // CK4-64: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 940 // CK4-32: [[TSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 941 // PRESENT=0x1000 | TARGET_PARAM=0x20 | TO=0x1 = 0x1021 942 // CK4: [[TTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1021]]] 943 944 // CK4-64: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 16] 945 // CK4-32: [[FSIZES:@.+]] = {{.+}}constant [1 x i64] [i64 8] 946 // PRESENT=0x1000 | TARGET_PARAM=0x20 | FROM=0x2 = 0x1022 947 // CK4: [[FTYPES:@.+]] = {{.+}}constant [1 x i64] [i64 [[#0x1022]]] 948 949 class C { 950 public: 951 int a; 952 double *b; 953 }; 954 955 #pragma omp declare mapper(id: C s) map(s.a, s.b[0:2]) 956 957 // CK4: define {{.*}}void [[MPRFUNC:@[.]omp_mapper[.].*C[.]id]](i8*{{.*}}, i8*{{.*}}, i8*{{.*}}, i64{{.*}}, i64{{.*}}) 958 // CK4: store i8* %{{[^,]+}}, i8** [[HANDLEADDR:%[^,]+]] 959 // CK4: store i8* %{{[^,]+}}, i8** [[BPTRADDR:%[^,]+]] 960 // CK4: store i8* %{{[^,]+}}, i8** [[VPTRADDR:%[^,]+]] 961 // CK4: store i64 %{{[^,]+}}, i{{64|32}}* [[SIZEADDR:%[^,]+]] 962 // CK4: store i64 %{{[^,]+}}, i64* [[TYPEADDR:%[^,]+]] 963 // CK4-DAG: [[BYTESIZE:%.+]] = load i64, i64* [[SIZEADDR]] 964 // CK4-64-DAG: [[SIZE:%.+]] = udiv exact i64 [[BYTESIZE]], 16 965 // CK4-32-DAG: [[SIZE:%.+]] = udiv exact i64 [[BYTESIZE]], 8 966 // CK4-DAG: [[TYPE:%.+]] = load i64, i64* [[TYPEADDR]] 967 // CK4-DAG: [[HANDLE:%.+]] = load i8*, i8** [[HANDLEADDR]] 968 // CK4-DAG: [[PTRBEGIN:%.+]] = bitcast i8** [[VPTRADDR]] to %class.C** 969 // CK4-DAG: [[PTREND:%.+]] = getelementptr %class.C*, %class.C** [[PTRBEGIN]], i64 [[SIZE]] 970 // CK4-DAG: [[BPTR:%.+]] = load i8*, i8** [[BPTRADDR]] 971 // CK4-DAG: [[BEGIN:%.+]] = load i8*, i8** [[VPTRADDR]] 972 // CK4: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 973 // CK4: br i1 [[ISARRAY]], label %[[INITEVALDEL:[^,]+]], label %[[LHEAD:[^,]+]] 974 975 // CK4: [[INITEVALDEL]] 976 // CK4: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 977 // CK4: [[ISNOTDEL:%.+]] = icmp eq i64 [[TYPEDEL]], 0 978 // CK4: br i1 [[ISNOTDEL]], label %[[INIT:[^,]+]], label %[[LHEAD:[^,]+]] 979 // CK4: [[INIT]] 980 // CK4-64-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 16 981 // CK4-32-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 8 982 // CK4-DAG: [[ITYPE:%.+]] = and i64 [[TYPE]], -4 983 // CK4: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[ITYPE]]) 984 // CK4: br label %[[LHEAD:[^,]+]] 985 986 // CK4: [[LHEAD]] 987 // CK4: [[ISEMPTY:%.+]] = icmp eq %class.C** [[PTRBEGIN]], [[PTREND]] 988 // CK4: br i1 [[ISEMPTY]], label %[[DONE:[^,]+]], label %[[LBODY:[^,]+]] 989 // CK4: [[LBODY]] 990 // CK4: [[PTR:%.+]] = phi %class.C** [ [[PTRBEGIN]], %[[LHEAD]] ], [ [[PTRNEXT:%.+]], %[[LCORRECT:[^,]+]] ] 991 // CK4: [[OBJ:%.+]] = load %class.C*, %class.C** [[PTR]] 992 // CK4-DAG: [[ABEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 0 993 // CK4-DAG: [[BBEGIN:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 1 994 // CK4-DAG: [[BBEGIN2:%.+]] = getelementptr inbounds %class.C, %class.C* [[OBJ]], i32 0, i32 1 995 // CK4-DAG: [[BARRBEGIN:%.+]] = load double*, double** [[BBEGIN2]] 996 // CK4-DAG: [[BARRBEGINGEP:%.+]] = getelementptr inbounds double, double* [[BARRBEGIN]], i[[sz:64|32]] 0 997 // CK4-DAG: [[BEND:%.+]] = getelementptr double*, double** [[BBEGIN]], i32 1 998 // CK4-DAG: [[ABEGINV:%.+]] = bitcast i32* [[ABEGIN]] to i8* 999 // CK4-DAG: [[BENDV:%.+]] = bitcast double** [[BEND]] to i8* 1000 // CK4-DAG: [[ABEGINI:%.+]] = ptrtoint i8* [[ABEGINV]] to i64 1001 // CK4-DAG: [[BENDI:%.+]] = ptrtoint i8* [[BENDV]] to i64 1002 // CK4-DAG: [[CSIZE:%.+]] = sub i64 [[BENDI]], [[ABEGINI]] 1003 // CK4-DAG: [[CUSIZE:%.+]] = sdiv exact i64 [[CSIZE]], ptrtoint (i8* getelementptr (i8, i8* null, i32 1) to i64) 1004 // CK4-DAG: [[BPTRADDR0BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 1005 // CK4-DAG: [[PTRADDR0BC:%.+]] = bitcast i32* [[ABEGIN]] to i8* 1006 // CK4-DAG: [[PRESIZE:%.+]] = call i64 @__tgt_mapper_num_components(i8* [[HANDLE]]) 1007 // CK4-DAG: [[SHIPRESIZE:%.+]] = shl i64 [[PRESIZE]], 48 1008 // CK4-DAG: br label %[[MEMBER:[^,]+]] 1009 // CK4-DAG: [[MEMBER]] 1010 // CK4-DAG: br i1 true, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 1011 // CK4-DAG: [[MEMBERCOM]] 1012 // CK4-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 32, [[SHIPRESIZE]] 1013 // CK4-DAG: br label %[[LTYPE]] 1014 // CK4-DAG: [[LTYPE]] 1015 // CK4-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 32, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 1016 // CK4-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 1017 // CK4-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 1018 // CK4-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 1019 // CK4-DAG: [[ALLOC]] 1020 // CK4-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 1021 // CK4-DAG: br label %[[TYEND:[^,]+]] 1022 // CK4-DAG: [[ALLOCELSE]] 1023 // CK4-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 1024 // CK4-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 1025 // CK4-DAG: [[TO]] 1026 // CK4-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 1027 // CK4-DAG: br label %[[TYEND]] 1028 // CK4-DAG: [[TOELSE]] 1029 // CK4-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 1030 // CK4-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 1031 // CK4-DAG: [[FROM]] 1032 // CK4-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 1033 // CK4-DAG: br label %[[TYEND]] 1034 // CK4-DAG: [[TYEND]] 1035 // CK4-DAG: [[PHITYPE0:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 1036 // CK4: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR0BC]], i8* [[PTRADDR0BC]], i64 [[CUSIZE]], i64 [[PHITYPE0]]) 1037 // CK4-DAG: [[BPTRADDR1BC:%.+]] = bitcast %class.C* [[OBJ]] to i8* 1038 // CK4-DAG: [[PTRADDR1BC:%.+]] = bitcast i32* [[ABEGIN]] to i8* 1039 // CK4-DAG: br label %[[MEMBER:[^,]+]] 1040 // CK4-DAG: [[MEMBER]] 1041 // CK4-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 1042 // CK4-DAG: [[MEMBERCOM]] 1043 // 281474976710659 == 0x1,000,000,003 1044 // CK4-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710659, [[SHIPRESIZE]] 1045 // CK4-DAG: br label %[[LTYPE]] 1046 // CK4-DAG: [[LTYPE]] 1047 // CK4-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710659, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 1048 // CK4-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 1049 // CK4-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 1050 // CK4-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 1051 // CK4-DAG: [[ALLOC]] 1052 // CK4-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 1053 // CK4-DAG: br label %[[TYEND:[^,]+]] 1054 // CK4-DAG: [[ALLOCELSE]] 1055 // CK4-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 1056 // CK4-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 1057 // CK4-DAG: [[TO]] 1058 // CK4-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 1059 // CK4-DAG: br label %[[TYEND]] 1060 // CK4-DAG: [[TOELSE]] 1061 // CK4-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 1062 // CK4-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 1063 // CK4-DAG: [[FROM]] 1064 // CK4-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 1065 // CK4-DAG: br label %[[TYEND]] 1066 // CK4-DAG: [[TYEND]] 1067 // CK4-DAG: [[TYPE1:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 1068 // CK4: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR1BC]], i8* [[PTRADDR1BC]], i64 4, i64 [[TYPE1]]) 1069 // CK4-DAG: [[BPTRADDR2BC:%.+]] = bitcast double** [[BBEGIN]] to i8* 1070 // CK4-DAG: [[PTRADDR2BC:%.+]] = bitcast double* [[BARRBEGINGEP]] to i8* 1071 // CK4-DAG: br label %[[MEMBER:[^,]+]] 1072 // CK4-DAG: [[MEMBER]] 1073 // CK4-DAG: br i1 false, label %[[LTYPE:[^,]+]], label %[[MEMBERCOM:[^,]+]] 1074 // CK4-DAG: [[MEMBERCOM]] 1075 // 281474976710675 == 0x1,000,000,013 1076 // CK4-DAG: [[MEMBERCOMTYPE:%.+]] = add nuw i64 281474976710675, [[SHIPRESIZE]] 1077 // CK4-DAG: br label %[[LTYPE]] 1078 // CK4-DAG: [[LTYPE]] 1079 // CK4-DAG: [[MEMBERTYPE:%.+]] = phi i64 [ 281474976710675, %[[MEMBER]] ], [ [[MEMBERCOMTYPE]], %[[MEMBERCOM]] ] 1080 // CK4-DAG: [[TYPETF:%.+]] = and i64 [[TYPE]], 3 1081 // CK4-DAG: [[ISALLOC:%.+]] = icmp eq i64 [[TYPETF]], 0 1082 // CK4-DAG: br i1 [[ISALLOC]], label %[[ALLOC:[^,]+]], label %[[ALLOCELSE:[^,]+]] 1083 // CK4-DAG: [[ALLOC]] 1084 // CK4-DAG: [[ALLOCTYPE:%.+]] = and i64 [[MEMBERTYPE]], -4 1085 // CK4-DAG: br label %[[TYEND:[^,]+]] 1086 // CK4-DAG: [[ALLOCELSE]] 1087 // CK4-DAG: [[ISTO:%.+]] = icmp eq i64 [[TYPETF]], 1 1088 // CK4-DAG: br i1 [[ISTO]], label %[[TO:[^,]+]], label %[[TOELSE:[^,]+]] 1089 // CK4-DAG: [[TO]] 1090 // CK4-DAG: [[TOTYPE:%.+]] = and i64 [[MEMBERTYPE]], -3 1091 // CK4-DAG: br label %[[TYEND]] 1092 // CK4-DAG: [[TOELSE]] 1093 // CK4-DAG: [[ISFROM:%.+]] = icmp eq i64 [[TYPETF]], 2 1094 // CK4-DAG: br i1 [[ISFROM]], label %[[FROM:[^,]+]], label %[[TYEND]] 1095 // CK4-DAG: [[FROM]] 1096 // CK4-DAG: [[FROMTYPE:%.+]] = and i64 [[MEMBERTYPE]], -2 1097 // CK4-DAG: br label %[[TYEND]] 1098 // CK4-DAG: [[TYEND]] 1099 // CK4-DAG: [[TYPE2:%.+]] = phi i64 [ [[ALLOCTYPE]], %[[ALLOC]] ], [ [[TOTYPE]], %[[TO]] ], [ [[FROMTYPE]], %[[FROM]] ], [ [[MEMBERTYPE]], %[[TOELSE]] ] 1100 // CK4: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTRADDR2BC]], i8* [[PTRADDR2BC]], i64 16, i64 [[TYPE2]]) 1101 // CK4: [[PTRNEXT]] = getelementptr %class.C*, %class.C** [[PTR]], i32 1 1102 // CK4: [[ISDONE:%.+]] = icmp eq %class.C** [[PTRNEXT]], [[PTREND]] 1103 // CK4: br i1 [[ISDONE]], label %[[LEXIT:[^,]+]], label %[[LBODY]] 1104 1105 // CK4: [[LEXIT]] 1106 // CK4: [[ISARRAY:%.+]] = icmp sge i64 [[SIZE]], 1 1107 // CK4: br i1 [[ISARRAY]], label %[[EVALDEL:[^,]+]], label %[[DONE]] 1108 // CK4: [[EVALDEL]] 1109 // CK4: [[TYPEDEL:%.+]] = and i64 [[TYPE]], 8 1110 // CK4: [[ISDEL:%.+]] = icmp ne i64 [[TYPEDEL]], 0 1111 // CK4: br i1 [[ISDEL]], label %[[DEL:[^,]+]], label %[[DONE]] 1112 // CK4: [[DEL]] 1113 // CK4-64-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 16 1114 // CK4-32-DAG: [[ARRSIZE:%.+]] = mul nuw i64 [[SIZE]], 8 1115 // CK4-DAG: [[DTYPE:%.+]] = and i64 [[TYPE]], -4 1116 // CK4: call void @__tgt_push_mapper_component(i8* [[HANDLE]], i8* [[BPTR]], i8* [[BEGIN]], i64 [[ARRSIZE]], i64 [[DTYPE]]) 1117 // CK4: br label %[[DONE]] 1118 // CK4: [[DONE]] 1119 // CK4: ret void 1120 1121 1122 // CK4-LABEL: define {{.*}}void @{{.*}}foo{{.*}} 1123 void foo(int a){ 1124 int i = a; 1125 C c; 1126 c.a = a; 1127 1128 // CK4-DAG: call void @__tgt_target_data_update_mapper(i64 -1, i32 1, i8** [[TGEPBP:%.+]], i8** [[TGEPP:%.+]], i64* getelementptr {{.+}}[1 x i64]* [[TSIZES]], i32 0, i32 0), {{.+}}getelementptr {{.+}}[1 x i64]* [[TTYPES]]{{.+}}, i8** [[TMPRGEP:%.+]]) 1129 // CK4-DAG: [[TGEPBP]] = getelementptr inbounds {{.+}}[[TBP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 1130 // CK4-DAG: [[TGEPP]] = getelementptr inbounds {{.+}}[[TP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 1131 // CK4-DAG: [[TMPRGEP]] = bitcast [1 x i8*]* [[TMPR:%[^,]+]] to i8** 1132 // CK4-DAG: [[TBP0:%.+]] = getelementptr inbounds {{.+}}[[TBP]], i{{.+}} 0, i{{.+}} 0 1133 // CK4-DAG: [[TP0:%.+]] = getelementptr inbounds {{.+}}[[TP]], i{{.+}} 0, i{{.+}} 0 1134 // CK4-DAG: [[TMPR1:%.+]] = getelementptr inbounds {{.+}}[[TMPR]], i[[sz]] 0, i[[sz]] 0 1135 // CK4-DAG: [[TCBP0:%.+]] = bitcast i8** [[TBP0]] to %class.C** 1136 // CK4-DAG: [[TCP0:%.+]] = bitcast i8** [[TP0]] to %class.C** 1137 // CK4-DAG: store %class.C* [[VAL:%[^,]+]], %class.C** [[TCBP0]] 1138 // CK4-DAG: store %class.C* [[VAL]], %class.C** [[TCP0]] 1139 // CK4-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[TMPR1]] 1140 #pragma omp target update to(present, mapper(id): c) 1141 1142 // CK4-DAG: call void @__tgt_target_data_update_mapper(i64 -1, i32 1, i8** [[FGEPBP:%.+]], i8** [[FGEPP:%.+]], i64* getelementptr {{.+}}[1 x i64]* [[FSIZES]], i32 0, i32 0), {{.+}}getelementptr {{.+}}[1 x i64]* [[FTYPES]]{{.+}}, i8** [[FMPRGEP:%.+]]) 1143 // CK4-DAG: [[FGEPBP]] = getelementptr inbounds {{.+}}[[FBP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 1144 // CK4-DAG: [[FGEPP]] = getelementptr inbounds {{.+}}[[FP:%[^,]+]], i{{.+}} 0, i{{.+}} 0 1145 // CK4-DAG: [[FMPRGEP]] = bitcast [1 x i8*]* [[FMPR:%[^,]+]] to i8** 1146 // CK4-DAG: [[FBP0:%.+]] = getelementptr inbounds {{.+}}[[FBP]], i{{.+}} 0, i{{.+}} 0 1147 // CK4-DAG: [[FP0:%.+]] = getelementptr inbounds {{.+}}[[FP]], i{{.+}} 0, i{{.+}} 0 1148 // CK4-DAG: [[FMPR1:%.+]] = getelementptr inbounds {{.+}}[[FMPR]], i[[sz]] 0, i[[sz]] 0 1149 // CK4-DAG: [[FCBP0:%.+]] = bitcast i8** [[FBP0]] to %class.C** 1150 // CK4-DAG: [[FCP0:%.+]] = bitcast i8** [[FP0]] to %class.C** 1151 // CK4-DAG: store %class.C* [[VAL]], %class.C** [[FCBP0]] 1152 // CK4-DAG: store %class.C* [[VAL]], %class.C** [[FCP0]] 1153 // CK4-DAG: store i8* bitcast (void (i8*, i8*, i8*, i64, i64)* [[MPRFUNC]] to i8*), i8** [[FMPR1]] 1154 #pragma omp target update from(mapper(id), present: c) 1155 } 1156 1157 #endif // CK4 1158 1159 #endif // HEADER 1160