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