summaryrefslogtreecommitdiffstats
diff options
context:
space:
mode:
Diffstat
-rw-r--r--clang/lib/CodeGen/CGStmtOpenMP.cpp42
-rw-r--r--clang/test/OpenMP/distribute_parallel_for_reduction_codegen.cpp84
2 files changed, 107 insertions, 19 deletions
diff --git a/clang/lib/CodeGen/CGStmtOpenMP.cpp b/clang/lib/CodeGen/CGStmtOpenMP.cpp
index b4cc7ad63ce..bf9a2572231 100644
--- a/clang/lib/CodeGen/CGStmtOpenMP.cpp
+++ b/clang/lib/CodeGen/CGStmtOpenMP.cpp
@@ -3358,26 +3358,30 @@ void CodeGenFunction::EmitOMPDistributeLoop(const OMPLoopDirective &S,
CGF.EmitLoadOfScalar(IL, S.getLocStart()));
});
}
- OpenMPDirectiveKind ReductionKind = OMPD_unknown;
- if (isOpenMPParallelDirective(S.getDirectiveKind()) &&
- isOpenMPSimdDirective(S.getDirectiveKind())) {
- ReductionKind = OMPD_parallel_for_simd;
- } else if (isOpenMPParallelDirective(S.getDirectiveKind())) {
- ReductionKind = OMPD_parallel_for;
- } else if (isOpenMPSimdDirective(S.getDirectiveKind())) {
- ReductionKind = OMPD_simd;
- } else if (!isOpenMPTeamsDirective(S.getDirectiveKind()) &&
- S.hasClausesOfKind<OMPReductionClause>()) {
- llvm_unreachable(
- "No reduction clauses is allowed in distribute directive.");
+ if (isOpenMPSimdDirective(S.getDirectiveKind()) &&
+ !isOpenMPParallelDirective(S.getDirectiveKind()) &&
+ !isOpenMPTeamsDirective(S.getDirectiveKind())) {
+ OpenMPDirectiveKind ReductionKind = OMPD_unknown;
+ if (isOpenMPParallelDirective(S.getDirectiveKind()) &&
+ isOpenMPSimdDirective(S.getDirectiveKind())) {
+ ReductionKind = OMPD_parallel_for_simd;
+ } else if (isOpenMPParallelDirective(S.getDirectiveKind())) {
+ ReductionKind = OMPD_parallel_for;
+ } else if (isOpenMPSimdDirective(S.getDirectiveKind())) {
+ ReductionKind = OMPD_simd;
+ } else if (!isOpenMPTeamsDirective(S.getDirectiveKind()) &&
+ S.hasClausesOfKind<OMPReductionClause>()) {
+ llvm_unreachable(
+ "No reduction clauses is allowed in distribute directive.");
+ }
+ EmitOMPReductionClauseFinal(S, ReductionKind);
+ // Emit post-update of the reduction variables if IsLastIter != 0.
+ emitPostUpdateForReductionClause(
+ *this, S, [&](CodeGenFunction &CGF) -> llvm::Value * {
+ return CGF.Builder.CreateIsNotNull(
+ CGF.EmitLoadOfScalar(IL, S.getLocStart()));
+ });
}
- EmitOMPReductionClauseFinal(S, ReductionKind);
- // Emit post-update of the reduction variables if IsLastIter != 0.
- emitPostUpdateForReductionClause(
- *this, S, [&](CodeGenFunction &CGF) -> llvm::Value * {
- return CGF.Builder.CreateIsNotNull(
- CGF.EmitLoadOfScalar(IL, S.getLocStart()));
- });
// Emit final copy of the lastprivate variables if IsLastIter != 0.
if (HasLastprivateClause) {
EmitOMPLastprivateClauseFinal(
diff --git a/clang/test/OpenMP/distribute_parallel_for_reduction_codegen.cpp b/clang/test/OpenMP/distribute_parallel_for_reduction_codegen.cpp
new file mode 100644
index 00000000000..dc2a3ca5350
--- /dev/null
+++ b/clang/test/OpenMP/distribute_parallel_for_reduction_codegen.cpp
@@ -0,0 +1,84 @@
+// Test host code gen
+
+// 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 --check-prefix CHECK --check-prefix CHECK-64
+// 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
+// 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 --check-prefix CHECK --check-prefix CHECK-64
+// RUN: %clang_cc1 -verify -fopenmp -fopenmp-version=45 -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
+// 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
+// RUN: %clang_cc1 -fopenmp -fopenmp-version=45 -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
+
+// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -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
+// 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
+// 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-ONLY1 %s
+// RUN: %clang_cc1 -verify -fopenmp-simd -fopenmp-version=45 -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
+// 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
+// RUN: %clang_cc1 -fopenmp-simd -fopenmp-version=45 -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
+// SIMD-ONLY1-NOT: {{__kmpc|__tgt}}
+// expected-no-diagnostics
+#ifndef HEADER
+#define HEADER
+
+
+template <typename T>
+T tmain(T &r) {
+ int n = 1000;
+ // schedule: dynamic chunk
+ #pragma omp target map(tofrom:r)
+ #pragma omp teams
+ #pragma omp distribute parallel for reduction(+:r)
+ for (int i = 0; i < n; ++i)
+ r += (T)i;
+
+ return r;
+}
+
+int main() {
+ int n = 1000;
+ int r = 0;
+ #pragma omp target map(tofrom:r)
+ #pragma omp teams
+ #pragma omp distribute parallel for reduction(+:r)
+ for (int i = 0; i < n; ++i)
+ r += i;
+
+ return tmain<int>(r);
+}
+
+// CHECK-LABEL: main
+// CHECK: call{{.+}} @__tgt_target_teams(
+// CHECK: call void [[OFFL:@.+]](
+// CHECK: call{{.+}} [[TMAIN:@.+]](i{{32|64}}
+// CHECK: ret
+
+// CHECK: define{{.+}} [[OFFL]](
+// CHECK: call{{.+}} @__kmpc_fork_teams({{.+}}, {{.+}}, {{.+}} [[TEOUTL:@.+]] to{{.+}}
+// CHECK: ret void
+
+// CHECK: define{{.+}} [[TEOUTL]](
+// CHECK: call{{.+}} @__kmpc_fork_call({{.+}}, {{.+}}, {{.+}} [[PAROUTL:@.+]] to{{.+}}
+// CHECK: ret void
+
+// CHECK: define{{.+}} [[PAROUTL]](
+// CHECK: call{{.+}} @__kmpc_reduce_nowait(
+// CHECK: call{{.+}} @__kmpc_end_reduce_nowait(
+// CHECK: ret void
+
+// CHECK: define{{.+}} [[TMAIN]](i{{32|64}}
+// CHECK: call{{.+}} @__tgt_target_teams(
+// CHECK: call void [[TOFFL:@.+]](
+// CHECK: ret
+
+// CHECK: define{{.+}} [[TOFFL]](
+// CHECK: call{{.+}} @__kmpc_fork_teams({{.+}}, {{.+}}, {{.+}} [[TEMPLTEOUTL:@.+]] to{{.+}}
+// CHECK: ret void
+
+// CHECK: define{{.+}} [[TEMPLTEOUTL]](
+// CHECK: call{{.+}} @__kmpc_fork_call({{.+}}, {{.+}}, {{.+}} [[TPAROUTL:@.+]] to{{.+}}
+// CHECK: ret void
+
+// CHECK: define{{.+}} [[TPAROUTL]](
+// CHECK: call{{.+}} @__kmpc_reduce_nowait(
+// CHECK: call{{.+}} @__kmpc_end_reduce_nowait(
+// CHECK: ret void
+
+#endif // HEADER
OpenPOWER on IntegriCloud