1 // Test host codegen only. 2 // 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 3 // RUN: %clang_cc1 -fopenmp -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 -x c++ -std=c++11 -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 5 // 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 6 // RUN: %clang_cc1 -fopenmp -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 -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 8 9 // expected-no-diagnostics 10 #ifndef HEADER 11 #define HEADER 12 13 // CHECK: [[ANON_T:%.+]] = type { ptr, ptr } 14 // CHECK-DAG: [[SIZES_TEMPLATE:@.+]] = private {{.+}} constant [5 x i[[PTRSZ:32|64]]] [i{{32|64}} 4, i{{32|64}} 4, i{{32|64}} {{8|16}}, i{{32|64}} 0, i{{32|64}} 0] 15 // CHECK-DAG: [[TYPES_TEMPLATE:@.+]] = private {{.+}} constant [5 x i64] [i64 800, i64 800, i64 673, i64 844424930132752, i64 844424930132752] 16 // CHECK-DAG: [[SIZES:@.+]] = private {{.+}} constant [3 x i[[PTRSZ:32|64]]] [i{{32|64}} {{8|16}}, i{{32|64}} 0, i{{32|64}} 0] 17 // CHECK-DAG: [[TYPES:@.+]] = private {{.+}} constant [3 x i64] [i64 673, i64 281474976711440, i64 281474976711440] 18 // CHECK-DAG: [[TYPES3:@.+]] = private {{.+}} constant [3 x i64] [i64 545, i64 281474976711440, i64 800] 19 // CHECK-DAG: [[TYPES11:@.+]] = private {{.+}} constant [5 x i64] [i64 800, i64 800, i64 549, i64 844424930132752, i64 844424930132752] 20 // CHECK-DAG: [[TYPES13:@.+]] = private {{.+}} constant [2 x i64] [i64 545, i64 281474976711440] 21 // CHECK-DAG: [[TYPES15:@.+]] = private {{.+}} constant [2 x i64] [i64 673, i64 281474976711440] 22 23 template <typename F> 24 void omp_loop(int start, int end, F body) { 25 #pragma omp target teams distribute parallel for 26 for (int i = start; i < end; ++i) { 27 body(i); 28 } 29 } 30 31 template <typename F> 32 void omp_loop_ref(int start, int end, F& body) { 33 #pragma omp target teams distribute parallel for map(always, to: body) 34 for (int i = start; i < end; ++i) { 35 body(i); 36 } 37 int *p; 38 const auto &body_ref = [=](int i) {p[i]=0;}; 39 #pragma omp target map(to: body_ref) 40 body_ref(10); 41 #pragma omp target 42 body_ref(10); 43 } 44 45 template <class FTy> 46 struct C { 47 static void xoo(const FTy& f) { 48 int x = 10; 49 #pragma omp target map(to:f) 50 f(x); 51 } 52 }; 53 54 template <class FTy> 55 void zoo(const FTy &functor) { 56 C<FTy>::xoo(functor); 57 } 58 59 // CHECK: define {{.*}}[[MAIN:@.+]]( 60 int main() 61 { 62 int* p = new int[100]; 63 int* q = new int[100]; 64 auto body = [=](int i){ 65 p[i] = q[i]; 66 }; 67 zoo([=](int i){p[i] = 0;}); 68 69 #pragma omp target teams distribute parallel for 70 for (int i = 0; i < 100; ++i) { 71 body(i); 72 } 73 74 // CHECK: [[BASE_PTRS:%.+]] = alloca [3 x ptr]{{.+}} 75 // CHECK: [[PTRS:%.+]] = alloca [3 x ptr]{{.+}} 76 77 // First gep of pointers inside lambdas to store the values across function call need to be ignored 78 // CHECK: {{%.+}} = getelementptr inbounds nuw [[ANON_T]], ptr %{{.+}}, i{{.+}} 0, i{{.+}} 0 79 // CHECK: {{%.+}} = getelementptr inbounds nuw [[ANON_T]], ptr %{{.+}}, i{{.+}} 0, i{{.+}} 1 80 81 // access of pointers inside lambdas 82 // CHECK: [[BASE_PTR1:%.+]] = getelementptr inbounds nuw [[ANON_T]], ptr %{{.+}}, i{{.+}} 0, i{{.+}} 0 83 // CHECK: [[PTR1:%.+]] = load ptr, ptr [[BASE_PTR1]] 84 // CHECK: [[BASE_PTR2:%.+]] = getelementptr inbounds nuw [[ANON_T]], ptr %{{.+}}, i{{.+}} 0, i{{.+}} 1 85 // CHECK: [[PTR2:%.+]] = load ptr, ptr [[BASE_PTR2]] 86 87 // storage of pointers in baseptrs and ptrs arrays 88 // CHECK: [[LOC_LAMBDA:%.+]] = getelementptr inbounds [3 x ptr], ptr [[BASE_PTRS]], i{{.+}} 0, i{{.+}} 0 89 // CHECK: store ptr %{{.+}}, ptr [[LOC_LAMBDA]]{{.+}} 90 // CHECK: [[LOC_LAMBDA:%.+]] = getelementptr inbounds [3 x ptr], ptr [[PTRS]], i{{.+}} 0, i{{.+}} 0 91 // CHECK: store ptr %{{.+}}, ptr [[LOC_LAMBDA]]{{.+}} 92 93 // CHECK: [[LOC_PTR1:%.+]] = getelementptr inbounds [3 x ptr], ptr [[BASE_PTRS]], i{{.+}} 0, i{{.+}} 1 94 // CHECK: store ptr [[BASE_PTR1]], ptr [[LOC_PTR1]]{{.+}} 95 // CHECK: [[LOC_PTR1:%.+]] = getelementptr inbounds [3 x ptr], ptr [[PTRS]], i{{.+}} 0, i{{.+}} 1 96 // CHECK: store ptr [[PTR1]], ptr [[LOC_PTR1]]{{.+}} 97 98 99 // CHECK: [[LOC_PTR2:%.+]] = getelementptr inbounds [3 x ptr], ptr [[BASE_PTRS]], i{{.+}} 0, i{{.+}} 2 100 // CHECK: store ptr [[BASE_PTR2]], ptr [[LOC_PTR2]]{{.+}} 101 // CHECK: [[LOC_PTR2:%.+]] = getelementptr inbounds [3 x ptr], ptr [[PTRS]], i{{.+}} 0, i{{.+}} 2 102 // CHECK: store ptr [[PTR2]], ptr [[LOC_PTR2]]{{.+}} 103 104 // actual target invocation 105 // CHECK: [[BASES_GEP:%.+]] = getelementptr {{.+}} [3 x ptr], ptr [[BASE_PTRS]], {{.+}} 0, {{.+}} 0 106 // CHECK: [[PTRS_GEP:%.+]] = getelementptr {{.+}} [3 x ptr], ptr [[PTRS]], {{.+}} 0, {{.+}} 0 107 // CHECK: {{%.+}} = call i32 @__tgt_target_kernel(ptr @{{.+}}, i64 -1, i32 0, i32 0, ptr @.{{.+}}.region_id, ptr %{{.+}}) 108 109 omp_loop(0,100,body); 110 omp_loop_ref(0,100,body); 111 } 112 113 // CHECK: [[BASE_PTRS:%.+]] = alloca [5 x ptr]{{.+}} 114 // CHECK: [[PTRS:%.+]] = alloca [5 x ptr]{{.+}} 115 116 // access of pointers inside lambdas 117 // CHECK: [[BASE_PTR1:%.+]] = getelementptr inbounds nuw [[ANON_T]], ptr %{{.+}}, i{{.+}} 0, i{{.+}} 0 118 // CHECK: [[PTR1:%.+]] = load ptr, ptr [[BASE_PTR1]] 119 // CHECK: [[BASE_PTR2:%.+]] = getelementptr inbounds nuw [[ANON_T]], ptr %{{.+}}, i{{.+}} 0, i{{.+}} 1 120 // CHECK: [[PTR2:%.+]] = load ptr, ptr [[BASE_PTR2]] 121 122 // storage of pointers in baseptrs and ptrs arrays 123 // CHECK: [[LOC_LAMBDA:%.+]] = getelementptr inbounds [5 x ptr], ptr [[BASE_PTRS]], i{{.+}} 0, i{{.+}} 2 124 // CHECK: store ptr %{{.+}}, ptr [[LOC_LAMBDA]]{{.+}} 125 // CHECK: [[LOC_LAMBDA:%.+]] = getelementptr inbounds [5 x ptr], ptr [[PTRS]], i{{.+}} 0, i{{.+}} 2 126 // CHECK: store ptr %{{.+}}, ptr [[LOC_LAMBDA]]{{.+}} 127 128 // CHECK: [[LOC_PTR1:%.+]] = getelementptr inbounds [5 x ptr], ptr [[BASE_PTRS]], i{{.+}} 0, i{{.+}} 3 129 // CHECK: store ptr [[BASE_PTR1]], ptr [[LOC_PTR1]]{{.+}} 130 // CHECK: [[LOC_PTR1:%.+]] = getelementptr inbounds [5 x ptr], ptr [[PTRS]], i{{.+}} 0, i{{.+}} 3 131 // CHECK: store ptr [[PTR1]], ptr [[LOC_PTR1]]{{.+}} 132 133 134 // CHECK: [[LOC_PTR2:%.+]] = getelementptr inbounds [5 x ptr], ptr [[BASE_PTRS]], i{{.+}} 0, i{{.+}} 4 135 // CHECK: store ptr [[BASE_PTR2]], ptr [[LOC_PTR2]]{{.+}} 136 // CHECK: [[LOC_PTR2:%.+]] = getelementptr inbounds [5 x ptr], ptr [[PTRS]], i{{.+}} 0, i{{.+}} 4 137 // CHECK: store ptr [[PTR2]], ptr [[LOC_PTR2]]{{.+}} 138 139 // actual target invocation 140 // CHECK: [[BASES_GEP:%.+]] = getelementptr {{.+}} [5 x ptr], ptr [[BASE_PTRS]], {{.+}} 0, {{.+}} 0 141 // CHECK: [[PTRS_GEP:%.+]] = getelementptr {{.+}} [5 x ptr], ptr [[PTRS]], {{.+}} 0, {{.+}} 0 142 143 // CHECK: define internal void @{{.+}}omp_loop_ref{{.+}}( 144 // CHECK: [[BODY:%body.addr]] = alloca ptr 145 // CHECK: [[TMP:%tmp]] = alloca ptr 146 // CHECK: [[BODY_REF:%body_ref]] = alloca ptr 147 // CHECK: [[REF_TMP:%ref.tmp]] = alloca %class.anon.1 148 // CHECK: [[TMP8:%tmp.+]] = alloca ptr 149 // CHECK: [[L0:%.+]] = load ptr, ptr [[BODY]] 150 // CHECK: store ptr [[L0]], ptr [[TMP]] 151 // CHECK: [[L5:%.+]] = load ptr, ptr [[TMP]] 152 // CHECK-NOT [[L6:%.+]] = load ptr, ptr [[TMP]] 153 // CHECK-NOT [[L7:%.+]] = load ptr, ptr [[TMP]] 154 // CHECK: store ptr [[REF_TMP]], ptr [[BODY_REF]] 155 // CHECK:[[L47:%.+]] = load ptr, ptr [[BODY_REF]] 156 // CHECK: store ptr [[L47]], ptr [[TMP8]] 157 // CHECK: [[L48:%.+]] = load ptr, ptr [[TMP8]] 158 // CHECK-NOT: [[L49:%.+]] = load ptr, ptr [[TMP8]] 159 // CHECK-NOT: [[L50:%.+]] = load ptr, ptr [[TMP8]] 160 // CHECK: ret void 161 162 // CHECK: define internal void @{{.+}}xoo{{.+}}( 163 // CHECK: [[FADDR:%f.addr]] = alloca ptr 164 // CHECK: [[L0:%.+]] = load ptr, ptr [[FADDR]] 165 // CHECK: store ptr [[L0]], ptr [[TMP:%tmp]] 166 // CHECK: [[L1:%.+]] = load ptr, ptr [[TMP]] 167 // CHECK-NOT: %4 = load ptr, ptr [[TMP]] 168 // CHECK-NOT: %5 = load ptr, ptr [[TMP]] 169 // CHECK: [[L4:%.+]] = getelementptr inbounds nuw %class.anon.0, ptr [[L1]], i32 0, i32 0 170 // CHECK: [[L5:%.+]] = load ptr, ptr [[L4]] 171 // CHECK: ret void 172 173 #endif 174