1 // Test host codegen. 2 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s 3 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 4 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s 5 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s 6 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s 7 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s 8 9 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 10 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 11 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 12 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 13 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s 14 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 15 // SIMD-ONLY0-NOT: {{__kmpc|__tgt}} 16 17 // Test target teams codegen - host bc file has to be created first. 18 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc 19 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s -check-prefix=TCHECK 20 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s 21 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s -check-prefix=TCHECK 22 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc 23 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck %s -check-prefix=TCHECK 24 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s 25 // RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s -check-prefix=TCHECK 26 27 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc 28 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck --check-prefix SIMD-ONLY1 %s 29 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s 30 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s 31 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc 32 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck --check-prefix SIMD-ONLY1 %s 33 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s 34 // RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s 35 // SIMD-ONLY1-NOT: {{__kmpc|__tgt}} 36 37 // Check that no target code is emitted if no omptests flag was provided. 38 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s -check-prefix=CHECK-NTARGET 39 40 // RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY2 %s 41 // SIMD-ONLY2-NOT: {{__kmpc|__tgt}} 42 43 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s 44 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 45 // RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s 46 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s 47 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s 48 // RUN: %clang_cc1 -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s 49 50 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 51 // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 52 // RUN: %clang_cc1 -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 53 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 54 // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s 55 // RUN: %clang_cc1 -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY0 %s 56 // SIMD-ONLY0-NOT: {{__kmpc|__tgt}} 57 58 // Test target teams codegen - host bc file has to be created first. 59 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc 60 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck %s -check-prefix=TCHECK 61 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s 62 // RUN: %clang_cc1 -fopenmp -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s -check-prefix=TCHECK 63 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc 64 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck %s -check-prefix=TCHECK 65 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s 66 // RUN: %clang_cc1 -fopenmp -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s -check-prefix=TCHECK 67 68 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm-bc %s -o %t-ppc-host.bc 69 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o - | FileCheck --check-prefix SIMD-ONLY1 %s 70 // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -o %t %s 71 // RUN: %clang_cc1 -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-ppc-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s 72 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm-bc %s -o %t-x86-host.bc 73 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o - | FileCheck --check-prefix SIMD-ONLY1 %s 74 // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -o %t %s 75 // RUN: %clang_cc1 -fopenmp-simd -x c++ -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -fopenmp-is-target-device -fopenmp-host-ir-file-path %t-x86-host.bc -include-pch %t -verify %s -emit-llvm -o - | FileCheck --check-prefix SIMD-ONLY1 %s 76 // SIMD-ONLY1-NOT: {{__kmpc|__tgt}} 77 78 // Check that no target code is emitted if no omptests flag was provided. 79 // RUN: %clang_cc1 -verify -fopenmp -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck %s -check-prefix=CHECK-NTARGET 80 81 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -triple powerpc64le-unknown-unknown -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY2 %s 82 // SIMD-ONLY2-NOT: {{__kmpc|__tgt}} 83 84 // expected-no-diagnostics 85 #ifndef HEADER 86 #define HEADER 87 88 // CHECK-DAG: [[SA:%.+]] = type { [4 x i32] } 89 // CHECK-DAG: [[SB:%.+]] = type { [8 x i32] } 90 // CHECK-DAG: [[SC:%.+]] = type { [16 x i32] } 91 // CHECK-DAG: [[SD:%.+]] = type { [32 x i32] } 92 // CHECK-DAG: [[SE:%.+]] = type { [64 x i32] } 93 // CHECK-DAG: [[ST1:%.+]] = type { [228 x i32] } 94 // CHECK-DAG: [[ST2:%.+]] = type { [1128 x i32] } 95 // CHECK-DAG: [[ENTTY:%.+]] = type { i64, i16, i16, i32, ptr, ptr, i64, i64, ptr } 96 97 // TCHECK: [[ENTTY:%.+]] = type { i64, i16, i16, i32, ptr, ptr, i64, i64, ptr } 98 99 // CHECK-DAG: [[A1:@.+]] = internal global [[SA]] 100 // CHECK-DAG: [[A2:@.+]] ={{.*}} global [[SA]] 101 // CHECK-DAG: [[B1:@.+]] ={{.*}} global [[SB]] 102 // CHECK-DAG: [[B2:@.+]] ={{.*}} global [[SB]] 103 // CHECK-DAG: [[C1:@.+]] = internal global [[SC]] 104 // CHECK-DAG: [[D1:@.+]] ={{.*}} global [[SD]] 105 // CHECK-DAG: [[E1:@.+]] ={{.*}} global [[SE]] 106 // CHECK-DAG: [[T1:@.+]] ={{.*}} global [[ST1]] 107 // CHECK-DAG: [[T2:@.+]] ={{.*}} global [[ST2]] 108 109 // CHECK-NTARGET-DAG: [[SA:%.+]] = type { [4 x i32] } 110 // CHECK-NTARGET-DAG: [[SB:%.+]] = type { [8 x i32] } 111 // CHECK-NTARGET-DAG: [[SC:%.+]] = type { [16 x i32] } 112 // CHECK-NTARGET-DAG: [[SD:%.+]] = type { [32 x i32] } 113 // CHECK-NTARGET-DAG: [[SE:%.+]] = type { [64 x i32] } 114 // CHECK-NTARGET-DAG: [[ST1:%.+]] = type { [228 x i32] } 115 // CHECK-NTARGET-DAG: [[ST2:%.+]] = type { [1128 x i32] } 116 // CHECK-NTARGET-NOT: type { ptr, ptr, % 117 // CHECK-NTARGET-NOT: type { i32, % 118 119 // We have 7 target regions 120 121 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 122 // TCHECK-NOT: {{@.+}} = weak{{.*}} constant i8 0 123 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 124 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 125 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 126 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 127 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 128 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 129 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 130 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 131 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 132 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 133 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 134 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 135 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 136 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 137 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 138 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 139 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 140 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 141 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 142 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 143 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 144 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 145 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 146 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 147 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 148 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 149 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 150 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 151 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 152 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 153 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 154 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 155 // CHECK-DAG: {{@.+}} = weak{{.*}} constant i8 0 156 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 4] 157 // CHECK-DAG: {{@.+}} = private unnamed_addr constant [1 x i64] [i64 800] 158 159 // CHECK-NTARGET-NOT: weak constant i8 0 160 // CHECK-NTARGET-NOT: private unnamed_addr constant [1 x i 161 162 // CHECK-DAG: [[NAMEPTR1:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME1:__omp_offloading_[0-9a-f]+_[0-9a-f]+__Z.+_l[0-9]+]]\00" 163 // CHECK-DAG: [[NAMEPTR2:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME2:.+]]\00" 164 // CHECK-DAG: [[NAMEPTR3:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME3:.+]]\00" 165 // CHECK-DAG: [[NAMEPTR4:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME4:.+]]\00" 166 // CHECK-DAG: [[NAMEPTR5:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME5:.+]]\00" 167 // CHECK-DAG: [[NAMEPTR6:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME6:.+]]\00" 168 // CHECK-DAG: [[NAMEPTR7:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME7:.+]]\00" 169 // CHECK-DAG: [[NAMEPTR8:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME8:.+]]\00" 170 // CHECK-DAG: [[NAMEPTR9:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME9:.+]]\00" 171 // CHECK-DAG: [[NAMEPTR10:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME10:.+]]\00" 172 // CHECK-DAG: [[NAMEPTR11:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME11:.+]]\00" 173 // CHECK-DAG: [[NAMEPTR12:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME12:.+]]\00" 174 175 // TCHECK-DAG: [[NAMEPTR1:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME1:__omp_offloading_[0-9a-f]+_[0-9a-f]+__Z.+_l[0-9]+]]\00" 176 // TCHECK-DAG: [[NAMEPTR2:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME2:.+]]\00" 177 // TCHECK-DAG: [[NAMEPTR3:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME3:.+]]\00" 178 // TCHECK-DAG: [[NAMEPTR4:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME4:.+]]\00" 179 // TCHECK-DAG: [[NAMEPTR5:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME5:.+]]\00" 180 // TCHECK-DAG: [[NAMEPTR6:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME6:.+]]\00" 181 // TCHECK-DAG: [[NAMEPTR7:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME7:.+]]\00" 182 // TCHECK-DAG: [[NAMEPTR8:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME8:.+]]\00" 183 // TCHECK-DAG: [[NAMEPTR9:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME9:.+]]\00" 184 // TCHECK-DAG: [[NAMEPTR10:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME10:.+]]\00" 185 // TCHECK-DAG: [[NAMEPTR11:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME11:.+]]\00" 186 // TCHECK-DAG: [[NAMEPTR12:@.+]] = internal unnamed_addr constant [{{.*}} x i8] c"[[NAME12:.+]]\00" 187 188 // We have 4 initializers, one for the 500 priority, another one for 501, or more for the default priority, and the last one for the offloading registration function. 189 // CHECK: @llvm.global_ctors = appending global [3 x { i32, ptr, ptr }] [ 190 // CHECK-SAME: { i32, ptr, ptr } { i32 500, ptr [[P500:@[^,]+]], ptr null }, 191 // CHECK-SAME: { i32, ptr, ptr } { i32 501, ptr [[P501:@[^,]+]], ptr null }, 192 // CHECK-SAME: { i32, ptr, ptr } { i32 65535, ptr [[PMAX:@[^,]+]], ptr null } 193 194 // CHECK-NTARGET: @llvm.global_ctors = appending global [3 x { i32, ptr, ptr }] [ 195 196 extern int *R; 197 198 struct SA { 199 int arr[4]; 200 void foo() { 201 int a = *R; 202 a += 1; 203 *R = a; 204 } 205 SA() { 206 int a = *R; 207 a += 2; 208 *R = a; 209 } 210 ~SA() { 211 int a = *R; 212 a += 3; 213 *R = a; 214 } 215 }; 216 217 struct SB { 218 int arr[8]; 219 void foo() { 220 int a = *R; 221 #pragma omp target teams 222 a += 4; 223 *R = a; 224 } 225 SB() { 226 int a = *R; 227 a += 5; 228 *R = a; 229 } 230 ~SB() { 231 int a = *R; 232 a += 6; 233 *R = a; 234 } 235 }; 236 237 struct SC { 238 int arr[16]; 239 void foo() { 240 int a = *R; 241 a += 7; 242 *R = a; 243 } 244 SC() { 245 int a = *R; 246 #pragma omp target teams 247 a += 8; 248 *R = a; 249 } 250 ~SC() { 251 int a = *R; 252 a += 9; 253 *R = a; 254 } 255 }; 256 257 struct SD { 258 int arr[32]; 259 void foo() { 260 int a = *R; 261 a += 10; 262 *R = a; 263 } 264 SD() { 265 int a = *R; 266 a += 11; 267 *R = a; 268 } 269 ~SD() { 270 int a = *R; 271 #pragma omp target teams 272 a += 12; 273 *R = a; 274 } 275 }; 276 277 struct SE { 278 int arr[64]; 279 void foo() { 280 int a = *R; 281 #pragma omp target teams if(target: 0) 282 a += 13; 283 *R = a; 284 } 285 SE() { 286 int a = *R; 287 #pragma omp target teams 288 a += 14; 289 *R = a; 290 } 291 ~SE() { 292 int a = *R; 293 #pragma omp target teams 294 a += 15; 295 *R = a; 296 } 297 }; 298 299 template <int x> 300 struct ST { 301 int arr[128 + x]; 302 void foo() { 303 int a = *R; 304 #pragma omp target teams 305 a += 16 + x; 306 *R = a; 307 } 308 ST() { 309 int a = *R; 310 #pragma omp target teams 311 a += 17 + x; 312 *R = a; 313 } 314 ~ST() { 315 int a = *R; 316 #pragma omp target teams 317 a += 18 + x; 318 *R = a; 319 } 320 }; 321 322 // We have to make sure we us all the target regions: 323 //CHECK-DAG: define internal void @[[NAME1]]( 324 //CHECK-DAG: call void @[[NAME1]]( 325 //CHECK-DAG: define internal void @[[NAME2]]( 326 //CHECK-DAG: call void @[[NAME2]]( 327 //CHECK-DAG: define internal void @[[NAME3]]( 328 //CHECK-DAG: call void @[[NAME3]]( 329 //CHECK-DAG: define internal void @[[NAME4]]( 330 //CHECK-DAG: call void @[[NAME4]]( 331 //CHECK-DAG: define internal void @[[NAME5]]( 332 //CHECK-DAG: call void @[[NAME5]]( 333 //CHECK-DAG: define internal void @[[NAME6]]( 334 //CHECK-DAG: call void @[[NAME6]]( 335 //CHECK-DAG: define internal void @[[NAME7]]( 336 //CHECK-DAG: call void @[[NAME7]]( 337 //CHECK-DAG: define internal void @[[NAME8]]( 338 //CHECK-DAG: call void @[[NAME8]]( 339 //CHECK-DAG: define internal void @[[NAME9]]( 340 //CHECK-DAG: call void @[[NAME9]]( 341 //CHECK-DAG: define internal void @[[NAME10]]( 342 //CHECK-DAG: call void @[[NAME10]]( 343 //CHECK-DAG: define internal void @[[NAME11]]( 344 //CHECK-DAG: call void @[[NAME11]]( 345 //CHECK-DAG: define internal void @[[NAME12]]( 346 //CHECK-DAG: call void @[[NAME12]]( 347 348 //TCHECK-DAG: define weak{{.*}} void @[[NAME1]]( 349 //TCHECK-DAG: define weak{{.*}} void @[[NAME2]]( 350 //TCHECK-DAG: define weak{{.*}} void @[[NAME3]]( 351 //TCHECK-DAG: define weak{{.*}} void @[[NAME4]]( 352 //TCHECK-DAG: define weak{{.*}} void @[[NAME5]]( 353 //TCHECK-DAG: define weak{{.*}} void @[[NAME6]]( 354 //TCHECK-DAG: define weak{{.*}} void @[[NAME7]]( 355 //TCHECK-DAG: define weak{{.*}} void @[[NAME8]]( 356 //TCHECK-DAG: define weak{{.*}} void @[[NAME9]]( 357 //TCHECK-DAG: define weak{{.*}} void @[[NAME10]]( 358 //TCHECK-DAG: define weak{{.*}} void @[[NAME11]]( 359 //TCHECK-DAG: define weak{{.*}} void @[[NAME12]]( 360 361 // CHECK-NTARGET-NOT: __tgt_target 362 363 // TCHECK-NOT: __tgt_target 364 365 // We have 2 initializers with priority 500 366 //CHECK: define internal void [[P500]]( 367 //CHECK: call void @{{.+}}() 368 //CHECK: call void @{{.+}}() 369 //CHECK-NOT: call void @{{.+}}() 370 //CHECK: ret void 371 372 // We have 1 initializers with priority 501 373 //CHECK: define internal void [[P501]]( 374 //CHECK: call void @{{.+}}() 375 //CHECK-NOT: call void @{{.+}}() 376 //CHECK: ret void 377 378 // We have 6 initializers with default priority 379 //CHECK: define internal void [[PMAX]]( 380 //CHECK: call void @{{.+}}() 381 //CHECK: call void @{{.+}}() 382 //CHECK: call void @{{.+}}() 383 //CHECK: call void @{{.+}}() 384 //CHECK: call void @{{.+}}() 385 //CHECK: call void @{{.+}}() 386 //CHECK-NOT: call void @{{.+}}() 387 //CHECK: ret void 388 389 static __attribute__((init_priority(500))) SA a1; 390 SA a2; 391 SB __attribute__((init_priority(500))) b1; 392 SB __attribute__((init_priority(501))) b2; 393 static SC c1; 394 SD d1; 395 SE e1; 396 ST<100> t1; 397 ST<1000> t2; 398 399 400 int bar(int a){ 401 int r = a; 402 403 a1.foo(); 404 a2.foo(); 405 b1.foo(); 406 b2.foo(); 407 c1.foo(); 408 d1.foo(); 409 e1.foo(); 410 t1.foo(); 411 t2.foo(); 412 413 #pragma omp target teams 414 ++r; 415 416 return r + *R; 417 } 418 419 // Check metadata is properly generated: 420 // CHECK: !omp_offload.info = !{!{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}} 421 // CHECK-DAG: = !{i32 0, i32 [[DEVID:-?[0-9]+]], i32 [[FILEID:-?[0-9]+]], !"_ZN2SB3fooEv", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 422 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SDD1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 423 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SEC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 424 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SED1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 425 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi1000EE3fooEv", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 426 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi100EEC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 427 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_Z3bari", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 428 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi100EED1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 429 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi1000EEC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 430 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi1000EED1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 431 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi100EE3fooEv", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 432 // CHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SCC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 433 434 // TCHECK: !omp_offload.info = !{!{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}, !{{[0-9]+}}} 435 // TCHECK-DAG: = !{i32 0, i32 [[DEVID:-?[0-9]+]], i32 [[FILEID:-?[0-9]+]], !"_ZN2SB3fooEv", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 436 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SDD1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 437 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SEC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 438 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SED1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 439 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi1000EE3fooEv", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 440 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi100EEC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 441 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_Z3bari", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 442 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi100EED1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 443 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi1000EEC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 444 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi1000EED1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 445 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2STILi100EE3fooEv", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 446 // TCHECK-DAG: = !{i32 0, i32 [[DEVID]], i32 [[FILEID]], !"_ZN2SCC1Ev", i32 {{[0-9]+}}, i32 0, i32 {{[0-9]+}}} 447 448 #endif 449