1 // add -fopenmp-targets 2 3 // RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-llvm %s -o - | FileCheck %s 4 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -emit-pch -o %t %s 5 // RUN: %clang_cc1 -fopenmp -x c++ -std=c++11 -triple powerpc64le-unknown-unknown -fopenmp-targets=powerpc64le-ibm-linux-gnu -include-pch %t -verify %s -emit-llvm -o - | FileCheck %s 6 // expected-no-diagnostics 7 #ifndef HEADER 8 #define HEADER 9 10 typedef __INTPTR_TYPE__ intptr_t; 11 12 // CHECK-DAG: [[IDENT_T_TY:%.+]] = type { i32, i32, i32, i32, i8* } 13 // CHECK-DAG: [[STR:@.+]] = private unnamed_addr constant [23 x i8] c";unknown;unknown;0;0;;\00" 14 // CHECK-DAG: [[DEF_LOC_2:@.+]] = private unnamed_addr constant [[IDENT_T_TY]] { i32 0, i32 2, i32 0, i32 0, i8* getelementptr inbounds ([23 x i8], [23 x i8]* [[STR]], i32 0, i32 0) } 15 16 void foo(); 17 18 struct S { 19 intptr_t a, b, c; 20 S(intptr_t a) : a(a) {} 21 operator char() { return a; } 22 ~S() {} 23 }; 24 25 template <typename T> 26 T tmain() { 27 #pragma omp target 28 #pragma omp teams 29 #pragma omp distribute parallel for proc_bind(master) 30 for(int i = 0; i < 1000; i++) {} 31 return T(); 32 } 33 34 int main() { 35 // CHECK-LABEL: @main 36 #pragma omp target 37 #pragma omp teams 38 #pragma omp distribute parallel for proc_bind(spread) 39 for(int i = 0; i < 1000; i++) {} 40 #pragma omp target 41 #pragma omp teams 42 #pragma omp distribute parallel for proc_bind(close) 43 for(int i = 0; i < 1000; i++) {} 44 return tmain<int>(); 45 } 46 47 // CHECK: call {{.*}}@__tgt_target_teams({{.+}}) 48 // CHECK: call void [[OFFL1:@.+]]() 49 // CHECK: call {{.*}}@__tgt_target_teams({{.+}}) 50 // CHECK: call void [[OFFL2:@.+]]() 51 // CHECK: [[CALL_RET:%.+]] = call{{.+}} i32 [[TMAIN:@.+]]() 52 // CHECK: ret i32 [[CALL_RET]] 53 54 // CHECK: define{{.+}} void [[OFFL1]]( 55 // CHECK: call {{.*}}void {{.+}} @__kmpc_fork_teams({{.+}}, {{.+}}, {{.+}}* [[OMP_OUTLINED_1:@.+]] to {{.+}}) 56 57 // CHECK: define{{.+}} [[OMP_OUTLINED_1]](i32* {{.+}} [[GTID_IN:%.+]], 58 // CHECK: [[GTID_ADDR:%.+]] = alloca i32*, 59 // CHECK: store i32* [[GTID_IN]], i32** [[GTID_ADDR]], 60 // CHECK: [[GTID_REF:%.+]] = load i32*, i32** [[GTID_ADDR]], 61 // CHECK: [[GTID_VAL:%.+]] = load i32, i32* [[GTID_REF]], 62 // CHECK: call {{.*}}void @__kmpc_push_proc_bind([[IDENT_T_TY]]* [[DEF_LOC_2]], i32 [[GTID_VAL]], i32 4) 63 // CHECK: call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call( 64 // CHECK: ret void 65 66 // CHECK: define{{.+}} [[OFFL2]]() 67 // CHECK: call {{.*}}void {{.+}} @__kmpc_fork_teams({{.+}}, {{.+}}, {{.+}}* [[OMP_OUTLINED_1:@.+]] to {{.+}}) 68 69 // CHECK: define{{.+}} [[OMP_OUTLINED_1]](i32* {{.+}} [[GTID_IN:%.+]], 70 // CHECK: [[GTID_ADDR:%.+]] = alloca i32*, 71 // CHECK: store i32* [[GTID_IN]], i32** [[GTID_ADDR]], 72 // CHECK: [[GTID_REF:%.+]] = load i32*, i32** [[GTID_ADDR]], 73 // CHECK: [[GTID_VAL:%.+]] = load i32, i32* [[GTID_REF]], 74 // CHECK: call {{.*}}void @__kmpc_push_proc_bind([[IDENT_T_TY]]* [[DEF_LOC_2]], i32 [[GTID_VAL]], i32 3) 75 // CHECK: call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call( 76 // CHECK: ret void 77 78 // CHECK: define{{.+}} [[TMAIN]]() 79 // CHECK: call {{.*}}@__tgt_target_teams({{.+}}) 80 // CHECK: call void [[OFFL3:@.+]]() 81 82 // CHECK: define{{.+}} [[OFFL3]]() 83 // CHECK: call {{.*}}void {{.+}} @__kmpc_fork_teams({{.+}}, {{.+}}, {{.+}}* [[OMP_OUTLINED_3:@.+]] to {{.+}}) 84 85 // CHECK: define{{.+}} [[OMP_OUTLINED_3]](i32* {{.+}} [[GTID_IN:%.+]], 86 // CHECK: [[GTID_ADDR:%.+]] = alloca i32*, 87 // CHECK: store i32* [[GTID_IN]], i32** [[GTID_ADDR]], 88 // CHECK: [[GTID_REF:%.+]] = load i32*, i32** [[GTID_ADDR]], 89 // CHECK: [[GTID_VAL:%.+]] = load i32, i32* [[GTID_REF]], 90 // CHECK: call {{.*}}void @__kmpc_push_proc_bind([[IDENT_T_TY]]* [[DEF_LOC_2]], i32 [[GTID_VAL]], i32 2) 91 // CHECK: call {{.*}}void (%ident_t*, i32, void (i32*, i32*, ...)*, ...) @__kmpc_fork_call( 92 // CHECK: ret void 93 #endif 94