262 lines · cpp
1// expected-no-diagnostics2#ifndef HEADER3#define HEADER4 5///==========================================================================///6// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-647// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s8// RUN: %clang_cc1 -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 CK27 --check-prefix CK27-649// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-3210// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s11// RUN: %clang_cc1 -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 CK27 --check-prefix CK27-3212 13// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-6414// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-version=45 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s15// RUN: %clang_cc1 -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 CK27 --check-prefix CK27-6416// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-3217// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-version=45 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s18// RUN: %clang_cc1 -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 CK27 --check-prefix CK27-3219 20// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-6421// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-version=50 -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s22// RUN: %clang_cc1 -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 CK27 --check-prefix CK27-6423// RUN: %clang_cc1 -DCK27 -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 CK27 --check-prefix CK27-3224// RUN: %clang_cc1 -DCK27 -fopenmp -fopenmp-version=50 -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s25// RUN: %clang_cc1 -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 CK27 --check-prefix CK27-3226 27// RUN: %clang_cc1 -DCK27 -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-ONLY26 %s28// RUN: %clang_cc1 -DCK27 -fopenmp-simd -fopenmp-targets=powerpc64le-ibm-linux-gnu -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -emit-pch -o %t %s29// RUN: %clang_cc1 -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-ONLY26 %s30// RUN: %clang_cc1 -DCK27 -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-ONLY26 %s31// RUN: %clang_cc1 -DCK27 -fopenmp-simd -fopenmp-targets=i386-pc-linux-gnu -x c++ -std=c++11 -triple i386-unknown-unknown -emit-pch -o %t %s32// RUN: %clang_cc1 -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-ONLY26 %s33// SIMD-ONLY26-NOT: {{__kmpc|__tgt}}34#ifdef CK2735 36// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 037// CK27: [[SIZE00:@.+]] = private {{.*}}constant [1 x i64] zeroinitializer38// CK27: [[MTYPE00:@.+]] = private {{.*}}constant [1 x i64] [i64 544]39 40// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 041// CK27: [[SIZE01:@.+]] = private {{.*}}constant [1 x i64] zeroinitializer42// CK27: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 35]43 44// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 045// CK27: [[SIZE02:@.+]] = private {{.*}}constant [1 x i64] zeroinitializer46// CK27: [[MTYPE02:@.+]] = private {{.*}}constant [1 x i64] [i64 35]47 48// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 049// CK27: [[SIZE03:@.+]] = private {{.*}}constant [1 x i64] zeroinitializer50// CK27: [[MTYPE03:@.+]] = private {{.*}}constant [1 x i64] [i64 35]51 52// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 053// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 054// CK27: [[SIZE05:@.+]] = private {{.*}}constant [1 x i64] zeroinitializer55// CK27: [[MTYPE05:@.+]] = private {{.*}}constant [1 x i64] [i64 32]56 57// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 058// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 059// CK27: [[SIZE07:@.+]] = private {{.*}}constant [1 x i64] [i64 4]60// CK27: [[MTYPE07:@.+]] = private {{.*}}constant [1 x i64] [i64 288]61 62// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 063// CK27-LABEL: @.__omp_offloading_{{.*}}zero_size_section_and_private_maps{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 064// CK27: [[SIZE09:@.+]] = private {{.*}}constant [1 x i64] [i64 40]65// CK27: [[MTYPE09:@.+]] = private {{.*}}constant [1 x i64] [i64 161]66 67// CK27-LABEL: zero_size_section_and_private_maps{{.*}}(68void zero_size_section_and_private_maps (int ii){69 70 // Map of a pointer.71 int *pa;72 73// Region 0074// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])75// CK27-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 276// CK27-DAG: store ptr [[BPGEP:%.+]], ptr [[BPARG]]77// CK27-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 378// CK27-DAG: store ptr [[PGEP:%.+]], ptr [[PARG]]79// CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]80// CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]81 82// CK27-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 083// CK27-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 084// CK27-DAG: store ptr [[VAR0:%.+]], ptr [[BP0]]85// CK27-DAG: store ptr [[VAR0]], ptr [[P0]]86 87// CK27: call void [[CALL00:@.+]](ptr {{[^,]+}})88#pragma omp target89 {90 pa[50]++;91 }92 93// Region 0194// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])95// CK27-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 296// CK27-DAG: store ptr [[BPGEP:%.+]], ptr [[BPARG]]97// CK27-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 398// CK27-DAG: store ptr [[PGEP:%.+]], ptr [[PARG]]99// CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]100// CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]101 102// CK27-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0103// CK27-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0104// CK27-DAG: store ptr [[RVAR0:%.+]], ptr [[BP0]]105// CK27-DAG: store ptr [[SEC0:%.+]], ptr [[P0]]106// CK27-DAG: [[RVAR0]] = load ptr, ptr [[VAR0:%[^,]+]]107// CK27-DAG: [[SEC0]] = getelementptr {{.*}}ptr [[RVAR00:%.+]], i{{.+}} 0108// CK27-DAG: [[RVAR00]] = load ptr, ptr [[VAR0]]109 110// CK27: call void [[CALL01:@.+]](ptr {{[^,]+}})111#pragma omp target map(pa[:0])112 {113 pa[50]++;114 }115 116// Region 02117// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])118// CK27-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2119// CK27-DAG: store ptr [[BPGEP:%.+]], ptr [[BPARG]]120// CK27-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3121// CK27-DAG: store ptr [[PGEP:%.+]], ptr [[PARG]]122// CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]123// CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]124 125// CK27-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0126// CK27-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0127// CK27-DAG: store ptr [[RVAR0:%.+]], ptr [[BP0]]128// CK27-DAG: store ptr [[SEC0:%.+]], ptr [[P0]]129// CK27-DAG: [[RVAR0]] = load ptr, ptr [[VAR0:%[^,]+]]130// CK27-DAG: [[SEC0]] = getelementptr {{.*}}ptr [[RVAR00:%.+]], i{{.+}} 0131// CK27-DAG: [[RVAR00]] = load ptr, ptr [[VAR0]]132 133// CK27: call void [[CALL02:@.+]](ptr {{[^,]+}})134#pragma omp target map(pa [0:0])135 {136 pa[50]++;137 }138 139// Region 03140// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])141// CK27-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2142// CK27-DAG: store ptr [[BPGEP:%.+]], ptr [[BPARG]]143// CK27-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3144// CK27-DAG: store ptr [[PGEP:%.+]], ptr [[PARG]]145// CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]146// CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]147 148// CK27-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0149// CK27-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0150// CK27-DAG: store ptr [[RVAR0:%.+]], ptr [[BP0]]151// CK27-DAG: store ptr [[SEC0:%.+]], ptr [[P0]]152// CK27-DAG: [[RVAR0]] = load ptr, ptr [[VAR0:%[^,]+]]153// CK27-DAG: [[SEC0]] = getelementptr {{.*}}ptr [[RVAR00:%.+]], i{{.+}} %{{.+}}154// CK27-DAG: [[RVAR00]] = load ptr, ptr [[VAR0]]155 156// CK27: call void [[CALL03:@.+]](ptr {{[^,]+}})157#pragma omp target map(pa [ii:0])158 {159 pa[50]++;160 }161 162 int *pvtPtr;163 int pvtScl;164 int pvtArr[10];165 166// Region 04167// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])168// CK27: call void [[CALL04:@.+]]()169#pragma omp target private(pvtPtr)170 {171 pvtPtr[5]++;172 }173 174// Region 05175// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])176// CK27-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2177// CK27-DAG: store ptr [[BPGEP:%.+]], ptr [[BPARG]]178// CK27-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3179// CK27-DAG: store ptr [[PGEP:%.+]], ptr [[PARG]]180// CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]181// CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]182 183// CK27-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0184// CK27-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0185// CK27-DAG: store ptr [[VAR0:%.+]], ptr [[BP0]]186// CK27-DAG: store ptr [[VAR0]], ptr [[P0]]187 188// CK27: call void [[CALL05:@.+]](ptr {{[^,]+}})189#pragma omp target firstprivate(pvtPtr)190 {191 pvtPtr[5]++;192 }193 194// Region 06195// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])196// CK27: call void [[CALL06:@.+]]()197#pragma omp target private(pvtScl)198 {199 pvtScl++;200 }201 202// Region 07203// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])204// CK27-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2205// CK27-DAG: store ptr [[BPGEP:%.+]], ptr [[BPARG]]206// CK27-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3207// CK27-DAG: store ptr [[PGEP:%.+]], ptr [[PARG]]208// CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BPS:%[^,]+]], i32 0, i32 0209// CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[PS:%[^,]+]], i32 0, i32 0210// CK27-DAG: [[BP1:%.+]] = getelementptr inbounds {{.+}}[[BPS]], i32 0, i32 0211// CK27-DAG: [[P1:%.+]] = getelementptr inbounds {{.+}}[[PS]], i32 0, i32 0212// CK27-DAG: store i[[Z:64|32]] [[VAL:%.+]], ptr [[BP1]]213// CK27-DAG: store i[[Z]] [[VAL]], ptr [[P1]]214// CK27-DAG: [[VAL]] = load i[[Z]], ptr [[ADDR:%.+]],215// CK27-64-DAG: store i32 {{.+}}, ptr [[ADDR]],216 217// CK27: call void [[CALL07:@.+]](i[[Z]] [[VAL]])218#pragma omp target firstprivate(pvtScl)219 {220 pvtScl++;221 }222 223// Region 08224// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])225// CK27: call void [[CALL08:@.+]]()226#pragma omp target private(pvtArr)227 {228 pvtArr[5]++;229 }230 231// Region 09232// CK27-DAG: call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 -1, i32 0, ptr @.{{.+}}.region_id, ptr [[ARGS:%.+]])233// CK27-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2234// CK27-DAG: store ptr [[BPGEP:%.+]], ptr [[BPARG]]235// CK27-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3236// CK27-DAG: store ptr [[PGEP:%.+]], ptr [[PARG]]237// CK27-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]]238// CK27-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]]239 240// CK27-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0241// CK27-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0242// CK27-DAG: store ptr [[VAR0:%.+]], ptr [[BP0]]243// CK27-DAG: store ptr [[VAR0]], ptr [[P0]]244 245// CK27: call void [[CALL09:@.+]](ptr {{[^,]+}})246#pragma omp target firstprivate(pvtArr)247 {248 pvtArr[5]++;249 }250}251 252// CK27: define {{.+}}[[CALL00]]253// CK27: define {{.+}}[[CALL01]]254// CK27: define {{.+}}[[CALL02]]255// CK27: define {{.+}}[[CALL03]]256// CK27: define {{.+}}[[CALL04]]257// CK27: define {{.+}}[[CALL05]]258// CK27: define {{.+}}[[CALL06]]259// CK27: define {{.+}}[[CALL07]]260#endif // CK27261#endif262