1 // expected-no-diagnostics 2 #ifndef HEADER 3 #define HEADER 4 5 ///==========================================================================/// 6 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-64 7 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-64 9 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-32 10 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-32 12 13 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-64 14 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-64 16 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-32 17 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-32 19 20 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-64 21 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-64 23 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-32 24 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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 CK31 --check-prefix CK31-32 26 27 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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-ONLY18 %s 28 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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-ONLY18 %s 30 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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-ONLY18 %s 31 // RUN: %clang_cc1 -no-opaque-pointers -DCK31 -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-ONLY18 %s 33 // SIMD-ONLY18-NOT: {{__kmpc|__tgt}} 34 #ifdef CK31 35 36 // CK31-LABEL: @.__omp_offloading_{{.*}}explicit_maps_single{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 37 // CK31: [[SIZE00:@.+]] = private {{.*}}constant [1 x i[[Z:64|32]]] [i[[Z:64|32]] 4] 38 // CK31: [[MTYPE00:@.+]] = private {{.*}}constant [1 x i64] [i64 1059] 39 40 // CK31-LABEL: @.__omp_offloading_{{.*}}explicit_maps_single{{.*}}_l{{[0-9]+}}.region_id = weak constant i8 0 41 // CK31: [[SIZE01:@.+]] = private {{.*}}constant [1 x i[[Z:64|32]]] [i[[Z:64|32]] 4] 42 // CK31: [[MTYPE01:@.+]] = private {{.*}}constant [1 x i64] [i64 1063] 43 44 // CK31-LABEL: explicit_maps_single{{.*}}( 45 void explicit_maps_single (int ii){ 46 // Map of a scalar. 47 int a = ii; 48 49 // Close. 50 // Region 00 51 // CK31-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]]) 52 // CK31-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2 53 // CK31-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]] 54 // CK31-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3 55 // CK31-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]] 56 // CK31-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] 57 // CK31-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] 58 59 // CK31-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 60 // CK31-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 61 // CK31-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to i32** 62 // CK31-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** 63 // CK31-DAG: store i32* [[VAR0:%.+]], i32** [[CBP0]] 64 // CK31-DAG: store i32* [[VAR0]], i32** [[CP0]] 65 66 // CK31: call void [[CALL00:@.+]](i32* {{[^,]+}}) 67 #pragma omp target map(close, tofrom \ 68 : a) 69 { 70 a++; 71 } 72 73 // Always Close. 74 // Region 01 75 // CK31-DAG: call i32 @__tgt_target_kernel(%struct.ident_t* @{{.+}}, i64 -1, i32 -1, i32 0, i8* @.{{.+}}.region_id, %struct.__tgt_kernel_arguments* [[ARGS:%.+]]) 76 // CK31-DAG: [[BPARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 2 77 // CK31-DAG: store i8** [[BPGEP:%.+]], i8*** [[BPARG]] 78 // CK31-DAG: [[PARG:%.+]] = getelementptr inbounds {{.+}}[[ARGS]], i32 0, i32 3 79 // CK31-DAG: store i8** [[PGEP:%.+]], i8*** [[PARG]] 80 // CK31-DAG: [[BPGEP]] = getelementptr inbounds {{.+}}[[BP:%[^,]+]] 81 // CK31-DAG: [[PGEP]] = getelementptr inbounds {{.+}}[[P:%[^,]+]] 82 83 // CK31-DAG: [[BP0:%.+]] = getelementptr inbounds {{.+}}[[BP]], i{{.+}} 0, i{{.+}} 0 84 // CK31-DAG: [[P0:%.+]] = getelementptr inbounds {{.+}}[[P]], i{{.+}} 0, i{{.+}} 0 85 // CK31-DAG: [[CBP0:%.+]] = bitcast i8** [[BP0]] to i32** 86 // CK31-DAG: [[CP0:%.+]] = bitcast i8** [[P0]] to i32** 87 // CK31-DAG: store i32* [[VAR0:%.+]], i32** [[CBP0]] 88 // CK31-DAG: store i32* [[VAR0]], i32** [[CP0]] 89 90 // CK31: call void [[CALL01:@.+]](i32* {{[^,]+}}) 91 #pragma omp target map(always close tofrom \ 92 : a) 93 { 94 a++; 95 } 96 } 97 // CK31: define {{.+}}[[CALL00]] 98 // CK31: define {{.+}}[[CALL01]] 99 100 #endif // CK31 101 #endif 102