1 // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix LAMBDA --check-prefix LAMBDA-64 2 // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 3 // RUN: %clang_cc1 -DLAMBDA -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 --check-prefix LAMBDA --check-prefix LAMBDA-64 4 // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix LAMBDA --check-prefix LAMBDA-32 5 // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s 6 // RUN: %clang_cc1 -DLAMBDA -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix LAMBDA --check-prefix LAMBDA-32 7 8 // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 9 // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 10 // RUN: %clang_cc1 -DLAMBDA -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 11 // RUN: %clang_cc1 -DLAMBDA -verify -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY0 %s 12 // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s 13 // RUN: %clang_cc1 -DLAMBDA -fopenmp-simd -x c++ -std=c++11 -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 14 // SIMD-ONLY0-NOT: {{__kmpc|__tgt}} 15 16 // RUN: %clang_cc1 -verify -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-64 17 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 18 // 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 --check-prefix CHECK --check-prefix CHECK-64 19 // RUN: %clang_cc1 -verify -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32 20 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-pch -o %t %s 21 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -std=c++11 -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s --check-prefix CHECK --check-prefix CHECK-32 22 23 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY1 %s 24 // 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 25 // 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-ONLY1 %s 26 // RUN: %clang_cc1 -verify -fopenmp-simd -x c++ -std=c++11 -triple i386-unknown-unknown -fopenmp-targets=i386-pc-linux-gnu -emit-llvm %s -o - | FileCheck --check-prefix SIMD-ONLY1 %s 27 // 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 28 // RUN: %clang_cc1 -fopenmp-simd -x c++ -std=c++11 -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-ONLY1 %s 29 // SIMD-ONLY1-NOT: {{__kmpc|__tgt}} 30 // expected-no-diagnostics 31 #ifndef HEADER 32 #define HEADER 33 34 template <class T> 35 struct S { 36 T f; 37 S(T a) : f(a) {} 38 S() : f() {} 39 operator T() { return T(); } 40 ~S() {} 41 }; 42 43 // CHECK: [[S_FLOAT_TY:%.+]] = type { float } 44 // CHECK: [[S_INT_TY:%.+]] = type { i{{[0-9]+}} } 45 template <typename T> 46 T tmain() { 47 S<T> test; 48 T t_var = T(); 49 T vec[] = {1, 2}; 50 S<T> s_arr[] = {1, 2}; 51 S<T> &var = test; 52 #pragma omp target 53 #pragma omp teams 54 #pragma omp distribute simd private(t_var, vec, s_arr, s_arr, var, var) 55 for (int i = 0; i < 2; ++i) { 56 vec[i] = t_var; 57 s_arr[i] = var; 58 } 59 return T(); 60 } 61 62 int main() { 63 static int svar; 64 volatile double g; 65 volatile double &g1 = g; 66 67 #ifdef LAMBDA 68 // LAMBDA-LABEL: @main 69 // LAMBDA: call{{.*}} void [[OUTER_LAMBDA:@.+]]( 70 [&]() { 71 static float sfvar; 72 // LAMBDA: define{{.*}} internal{{.*}} void [[OUTER_LAMBDA]]( 73 // LAMBDA: call i{{[0-9]+}} @__tgt_target_teams( 74 // LAMBDA: call void [[OFFLOADING_FUN:@.+]]( 75 76 // LAMBDA: define{{.+}} void [[OFFLOADING_FUN]]() 77 // LAMBDA: call {{.*}}void {{.+}} @__kmpc_fork_teams({{.+}}, i32 0, {{.+}}* [[OMP_OUTLINED:@.+]] to {{.+}}) 78 #pragma omp target 79 #pragma omp teams 80 #pragma omp distribute simd private(g, g1, svar, sfvar) 81 for (int i = 0; i < 2; ++i) { 82 // LAMBDA: define{{.*}} internal{{.*}} void [[OMP_OUTLINED]](i32* noalias %{{.+}}, i32* noalias %{{.+}}) 83 // LAMBDA: [[G_PRIVATE_ADDR:%.+]] = alloca double, 84 // LAMBDA: [[G1_PRIVATE_ADDR:%.+]] = alloca double, 85 // LAMBDA: [[TMP_PRIVATE_ADDR:%.+]] = alloca double*, 86 // LAMBDA: [[SVAR_PRIVATE_ADDR:%.+]] = alloca i{{[0-9]+}}, 87 // LAMBDA: [[SFVAR_PRIVATE_ADDR:%.+]] = alloca float, 88 // LAMBDA: store double* [[G1_PRIVATE_ADDR]], double** [[TMP_PRIVATE_ADDR]], 89 g = 1; 90 g1 = 1; 91 svar = 3; 92 sfvar = 4.0; 93 // LAMBDA: call {{.*}}void @__kmpc_for_static_init_4( 94 // LAMBDA: store double 1.0{{.+}}, double* [[G_PRIVATE_ADDR]], 95 // LAMBDA: store i{{[0-9]+}} 3, i{{[0-9]+}}* [[SVAR_PRIVATE_ADDR]], 96 // LAMBDA: store float 4.0{{.+}}, float* [[SFVAR_PRIVATE_ADDR]], 97 // LAMBDA: [[G_PRIVATE_ADDR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG:%.+]], i{{[0-9]+}} 0, i{{[0-9]+}} 0 98 // LAMBDA: store double* [[G_PRIVATE_ADDR]], double** [[G_PRIVATE_ADDR_REF]], 99 // LAMBDA: [[TMP_PRIVATE_ADDR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG:%.+]], i{{[0-9]+}} 0, i{{[0-9]+}} 1 100 // LAMBDA: [[G1_PRIVATE_ADDR_FROM_TMP:%.+]] = load double*, double** [[TMP_PRIVATE_ADDR]], 101 // LAMBDA: store double* [[G1_PRIVATE_ADDR_FROM_TMP]], double** [[TMP_PRIVATE_ADDR_REF]], 102 // LAMBDA: [[SVAR_PRIVATE_ADDR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG:%.+]], i{{[0-9]+}} 0, i{{[0-9]+}} 2 103 // LAMBDA: store i{{[0-9]+}}* [[SVAR_PRIVATE_ADDR]], i{{[0-9]+}}** [[SVAR_PRIVATE_ADDR_REF]] 104 // LAMBDA: [[SFVAR_PRIVATE_ADDR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG:%.+]], i{{[0-9]+}} 0, i{{[0-9]+}} 3 105 // LAMBDA: store float* [[SFVAR_PRIVATE_ADDR]], float** [[SFVAR_PRIVATE_ADDR_REF]] 106 // LAMBDA: call{{.*}} void [[INNER_LAMBDA:@.+]](%{{.+}}* [[ARG]]) 107 // LAMBDA: call {{.*}}void @__kmpc_for_static_fini( 108 [&]() { 109 // LAMBDA: define {{.+}} void [[INNER_LAMBDA]](%{{.+}}* [[ARG_PTR:%.+]]) 110 // LAMBDA: store %{{.+}}* [[ARG_PTR]], %{{.+}}** [[ARG_PTR_REF:%.+]], 111 g = 2; 112 g1 = 2; 113 svar = 4; 114 sfvar = 8.0; 115 // LAMBDA: [[ARG_PTR:%.+]] = load %{{.+}}*, %{{.+}}** [[ARG_PTR_REF]] 116 // LAMBDA: [[G_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 0 117 // LAMBDA: [[G_REF:%.+]] = load double*, double** [[G_PTR_REF]] 118 // LAMBDA: store double 2.0{{.+}}, double* [[G_REF]] 119 120 // LAMBDA: [[TMP_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 1 121 // LAMBDA: [[G1_REF:%.+]] = load double*, double** [[TMP_PTR_REF]] 122 // LAMBDA: store double 2.0{{.+}}, double* [[G1_REF]], 123 // LAMBDA: [[SVAR_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 2 124 // LAMBDA: [[SVAR_REF:%.+]] = load i{{[0-9]+}}*, i{{[0-9]+}}** [[SVAR_PTR_REF]] 125 // LAMBDA: store i{{[0-9]+}} 4, i{{[0-9]+}}* [[SVAR_REF]] 126 // LAMBDA: [[SFVAR_PTR_REF:%.+]] = getelementptr inbounds %{{.+}}, %{{.+}}* [[ARG_PTR]], i{{[0-9]+}} 0, i{{[0-9]+}} 3 127 // LAMBDA: [[SFVAR_REF:%.+]] = load float*, float** [[SFVAR_PTR_REF]] 128 // LAMBDA: store float 8.0{{.+}}, float* [[SFVAR_REF]] 129 }(); 130 } 131 }(); 132 return 0; 133 #else 134 S<float> test; 135 int t_var = 0; 136 int vec[] = {1, 2}; 137 S<float> s_arr[] = {1, 2}; 138 S<float> &var = test; 139 140 #pragma omp target 141 #pragma omp teams 142 #pragma omp distribute simd private(t_var, vec, s_arr, s_arr, var, var, svar) 143 for (int i = 0; i < 2; ++i) { 144 vec[i] = t_var; 145 s_arr[i] = var; 146 } 147 int i; 148 149 #pragma omp target 150 #pragma omp teams 151 #pragma omp distribute simd 152 for (i = 0; i < 2; ++i) { 153 ; 154 } 155 return tmain<int>(); 156 #endif 157 } 158 159 // CHECK: define{{.*}} i{{[0-9]+}} @main() 160 // CHECK: [[TEST:%.+]] = alloca [[S_FLOAT_TY]], 161 // CHECK: call {{.*}} [[S_FLOAT_TY_DEF_CONSTR:@.+]]([[S_FLOAT_TY]]* [[TEST]]) 162 // CHECK: call i{{[0-9]+}} @__tgt_target_teams( 163 // CHECK: call void [[OFFLOAD_FUN:@.+]]( 164 // CHECK: ret 165 166 // CHECK: define{{.+}} [[OFFLOAD_FUN]]() 167 // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 0, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*)* [[OMP_OUTLINED:@.+]] to void 168 // CHECK: ret 169 // 170 // CHECK: define internal void [[OMP_OUTLINED]](i{{[0-9]+}}* noalias [[GTID_ADDR:%.+]], i{{[0-9]+}}* noalias %{{.+}}) 171 // CHECK: [[T_VAR_PRIV:%.+]] = alloca i{{[0-9]+}}, 172 // CHECK: [[VEC_PRIV:%.+]] = alloca [2 x i{{[0-9]+}}], 173 // CHECK: [[S_ARR_PRIV:%.+]] = alloca [2 x [[S_FLOAT_TY]]], 174 // CHECK-NOT: alloca [2 x [[S_FLOAT_TY]]], 175 // CHECK: [[VAR_PRIV:%.+]] = alloca [[S_FLOAT_TY]], 176 // CHECK-NOT: alloca [[S_FLOAT_TY]], 177 // CHECK: [[S_VAR_PRIV:%.+]] = alloca i{{[0-9]+}}, 178 // CHECK: store i{{[0-9]+}}* [[GTID_ADDR]], i{{[0-9]+}}** [[GTID_ADDR_REF:%.+]] 179 // CHECK-NOT: [[T_VAR_PRIV]] 180 // CHECK-NOT: [[VEC_PRIV]] 181 // CHECK: {{.+}}: 182 // CHECK: [[S_ARR_PRIV_ITEM:%.+]] = phi [[S_FLOAT_TY]]* 183 // CHECK: call {{.*}} [[S_FLOAT_TY_DEF_CONSTR]]([[S_FLOAT_TY]]* [[S_ARR_PRIV_ITEM]]) 184 // CHECK-NOT: [[T_VAR_PRIV]] 185 // CHECK-NOT: [[VEC_PRIV]] 186 // CHECK: call {{.*}} [[S_FLOAT_TY_DEF_CONSTR]]([[S_FLOAT_TY]]* [[VAR_PRIV]]) 187 // CHECK: call void @__kmpc_for_static_init_4( 188 // CHECK: call void @__kmpc_for_static_fini( 189 // CHECK: ret void 190 191 // CHECK: define{{.*}} i{{[0-9]+}} [[TMAIN_INT:@.+]]() 192 // CHECK: [[TEST:%.+]] = alloca [[S_INT_TY]], 193 // CHECK: call {{.*}} [[S_INT_TY_DEF_CONSTR:@.+]]([[S_INT_TY]]* [[TEST]]) 194 // CHECK: call i{{[0-9]+}} @__tgt_target_teams( 195 // CHECK: call void [[OFFLOAD_FUN_1:@.+]]( 196 // CHECK: ret 197 198 199 // CHECK: define internal void [[OFFLOAD_FUN_1]]() 200 // CHECK: call void (%{{.+}}*, i{{[0-9]+}}, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)*, ...) @__kmpc_fork_teams(%{{.+}}* @{{.+}}, i{{[0-9]+}} 0, void (i{{[0-9]+}}*, i{{[0-9]+}}*, ...)* bitcast (void (i{{[0-9]+}}*, i{{[0-9]+}}*)* [[OMP_OUTLINED_1:@.+]] to void 201 // CHECK: ret 202 // 203 // CHECK: define internal void [[OMP_OUTLINED_1]](i{{[0-9]+}}* noalias [[GTID_ADDR:%.+]], i{{[0-9]+}}* noalias %{{.+}}) 204 // CHECK: [[T_VAR_PRIV:%.+]] = alloca i{{[0-9]+}}, 205 // CHECK: [[VEC_PRIV:%.+]] = alloca [2 x i{{[0-9]+}}], 206 // CHECK: [[S_ARR_PRIV:%.+]] = alloca [2 x [[S_INT_TY]]], 207 // CHECK-NOT: alloca [2 x [[S_INT_TY]]], 208 // CHECK: [[VAR_PRIV:%.+]] = alloca [[S_INT_TY]], 209 // CHECK-NOT: alloca [[S_INT_TY]], 210 // CHECK: store i{{[0-9]+}}* [[GTID_ADDR]], i{{[0-9]+}}** [[GTID_ADDR_REF:%.+]] 211 // CHECK-NOT: [[T_VAR_PRIV]] 212 // CHECK-NOT: [[VEC_PRIV]] 213 // CHECK: {{.+}}: 214 // CHECK: [[S_ARR_PRIV_ITEM:%.+]] = phi [[S_INT_TY]]* 215 // CHECK: call {{.*}} [[S_INT_TY_DEF_CONSTR]]([[S_INT_TY]]* [[S_ARR_PRIV_ITEM]]) 216 // CHECK-NOT: [[T_VAR_PRIV]] 217 // CHECK-NOT: [[VEC_PRIV]] 218 // CHECK: call {{.*}} [[S_INT_TY_DEF_CONSTR]]([[S_INT_TY]]* [[VAR_PRIV]]) 219 // CHECK: call void @__kmpc_for_static_init_4( 220 // CHECK: call void @__kmpc_for_static_fini( 221 // CHECK: ret void 222 223 // CHECK: !{!"llvm.loop.vectorize.enable", i1 true} 224 #endif 225