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