koops updated this revision to Diff 514096.
koops added a comment.
1. Adding semantic test clang/test/OpenMP/loop_bind_messages.cpp.
2. Changes suggested by Alexey.
3. >Why need to drop bind clause here? The new Directives to which loop
directive is being mapped to, do not contain the bind clause. a) omp loop
bind(parallel) ----> omp for b) omp loop bind(teams) -----> omp distribute
c) omp loop bind(thread) ------> omp simd
CHANGES SINCE LAST ACTION
https://reviews.llvm.org/D144634/new/
https://reviews.llvm.org/D144634
Files:
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/lib/Sema/SemaOpenMP.cpp
clang/test/OpenMP/loop_bind_codegen.cpp
clang/test/OpenMP/loop_bind_messages.cpp
Index: clang/test/OpenMP/loop_bind_messages.cpp
===================================================================
--- /dev/null
+++ clang/test/OpenMP/loop_bind_messages.cpp
@@ -0,0 +1,76 @@
+#ifndef HEADER
+#define HEADER
+// RUN: %clang_cc1 -triple x86_64-pc-linux-gnu -fopenmp -fopenmp-version=50 -verify %s
+
+#define NNN 50
+int aaa[NNN];
+
+void parallel_loop() {
+ #pragma omp parallel
+ {
+ #pragma omp loop
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+ }
+}
+
+void teams_loop() {
+ int var1, var2;
+
+ #pragma omp teams
+ {
+ #pragma omp loop bind(teams)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+
+ #pragma omp loop bind(teams) collapse(2) private(var1)
+ for (int i = 0 ; i < 3 ; i++) {
+ for (int j = 0 ; j < NNN ; j++) {
+ var1 += aaa[j];
+ }
+ }
+ }
+}
+
+void orphan_loop_with_bind() {
+ #pragma omp loop bind(parallel)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+}
+
+void orphan_loop_no_bind() {
+ #pragma omp loop // expected-error{{expected 'bind' clause for loop construct without an enclosing OpenMP construct}}
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+}
+
+void teams_loop_reduction() {
+ int total = 0;
+
+ #pragma omp teams
+ {
+ #pragma omp loop bind(teams)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+
+ #pragma omp loop bind(teams) reduction(+:total) // expected-error{{reduction clause not handled with '#pragma omp loop bind(teams)'}}
+ for (int j = 0 ; j < NNN ; j++) {
+ total+=aaa[j];
+ }
+ }
+}
+
+int main(int argc, char *argv[]) {
+ parallel_loop();
+ teams_loop();
+ orphan_loop_with_bind();
+ orphan_loop_no_bind();
+ teams_loop_reduction();
+}
+
+#endif
Index: clang/test/OpenMP/loop_bind_codegen.cpp
===================================================================
--- /dev/null
+++ clang/test/OpenMP/loop_bind_codegen.cpp
@@ -0,0 +1,151 @@
+// expected-no-diagnostics
+#ifndef HEADER
+#define HEADER
+// RUN: %clang_cc1 -DCK1 -verify -fopenmp -x c++ %s -emit-llvm -o - | FileCheck %s
+// RUN: %clang_cc1 -verify -fopenmp -x c++ -triple x86_64-unknown-unknown -emit-llvm %s -o - | FileCheck %s
+
+
+/*
+#include <stdio.h>
+#include <assert.h>
+#include <pthread.h>
+#include <omp.h>
+*/
+
+#define NNN 50
+int aaa[NNN];
+
+void parallel_loop() {
+ #pragma omp parallel
+ {
+ #pragma omp loop bind(parallel)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+ }
+}
+
+void parallel_loop_orphan() {
+ #pragma omp loop bind(parallel)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+}
+
+
+void teams_loop() {
+ #pragma omp teams
+ {
+ #pragma omp loop bind(teams)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+ }
+}
+
+void thread_loop() {
+ #pragma omp parallel
+ {
+ #pragma omp loop bind(thread)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+ }
+}
+
+void thread_loop_orphan() {
+ #pragma omp loop bind(thread)
+ for (int j = 0 ; j < NNN ; j++) {
+ aaa[j] = j*NNN;
+ }
+}
+
+int main() {
+ parallel_loop();
+ parallel_loop_orphan();
+ teams_loop();
+ thread_loop();
+ thread_loop_orphan();
+
+ return 0;
+}
+#endif
+// CHECK-LABEL: define {{[^@]+}}@_Z13parallel_loopv
+// CHECK-NEXT: entry:
+// CHECK-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @3, i32 0, ptr @.omp_outlined.)
+// CHECK-LABEL: define {{[^@]+}}@.omp_outlined
+// CHECK-NEXT: entry:
+// CHECK-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8
+// CHECK-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8
+// CHECK-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[TMP:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[J:%.*]] = alloca i32, align 4
+// CHECK: call void @__kmpc_for_static_init_4(ptr @1, i32 %1, i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
+// CHECK-LABEL: cond.true:
+// CHECK-NEXT: br label [[COND_END:%.*]]
+// CHECK-LABEL: cond.false:
+// CHECK-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
+// CHECK-NEXT: br label [[COND_END]]
+// CHECK: omp.inner.for.cond:
+// CHECK-NEXT: [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
+// CHECK: omp.inner.for.body:
+// CHECK: omp.loop.exit:
+// CHECK-NEXT: call void @__kmpc_for_static_fini(ptr @1, i32 %1)
+// CHECK-NEXT: call void @__kmpc_barrier(ptr @2, i32 %1)
+// CHECK-NEXT: ret void
+//
+// CHECK-LABEL: define {{[^@]+}}@_Z20parallel_loop_orphanv
+// CHECK-NEXT: entry:
+// CHECK-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4
+// CHECK: call void @__kmpc_for_static_init_4(ptr @1, i32 %0, i32 34, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
+// CHECK-LABEL: cond.true:
+// CHECK: omp.inner.for.cond:
+// CHECK: omp.loop.exit:
+//
+// CHECK-LABEL: define {{[^@]+}}@_Z10teams_loopv
+// CHECK-NEXT: entry:
+// CHECK-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_teams(ptr @3, i32 0, ptr @.omp_outlined..1)
+// CHECK-NEXT: ret void
+//
+// CHECK-LABEL: define {{[^@]+}}@.omp_outlined
+// CHECK-NEXT: entry:
+// CHECK-NEXT: [[DOTGLOBAL_TID__ADDR:%.*]] = alloca ptr, align 8
+// CHECK-NEXT: [[DOTBOUND_TID__ADDR:%.*]] = alloca ptr, align 8
+// CHECK-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[TMP:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_LB:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_UB:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_STRIDE:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_IS_LAST:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[J:%.*]] = alloca i32, align 4
+// CHECK: call void @__kmpc_for_static_init_4(ptr @4, i32 %1, i32 92, ptr [[DOTOMP_IS_LAST]], ptr [[DOTOMP_LB]], ptr [[DOTOMP_UB]], ptr [[DOTOMP_STRIDE]], i32 1, i32 1)
+// CHECK-LABEL: cond.true:
+// CHECK-NEXT: br label [[COND_END:%.*]]
+// CHECK-LABEL: cond.false:
+// CHECK-NEXT: [[TMP11:%.*]] = load i32, ptr [[DOTOMP_UB]], align 4
+// CHECK-NEXT: br label [[COND_END]]
+// CHECK: omp.inner.for.cond:
+// CHECK-NEXT: [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
+// CHECK: omp.inner.for.body:
+// CHECK: omp.loop.exit:
+// CHECK-NEXT: call void @__kmpc_for_static_fini(ptr @4, i32 %1)
+// CHECK-NEXT: ret void
+//
+// CHECK-LABEL: define {{[^@]+}}@_Z11thread_loopv
+// CHECK-NEXT: entry:
+// CHECK-NEXT: call void (ptr, i32, ptr, ...) @__kmpc_fork_call(ptr @3, i32 0, ptr @.omp_outlined..2)
+// CHECK-LABEL: define {{[^@]+}}@.omp_outlined
+// CHECK: omp.inner.for.cond:
+//
+// CHECK-LABEL: define {{[^@]+}}@_Z18thread_loop_orphanv
+// CHECK-NEXT: entry:
+// CHECK-NEXT: [[TMP:%.*]] = alloca i32, align 4
+// CHECK-NEXT: [[DOTOMP_IV:%.*]] = alloca i32, align 4
+// CHECK: omp.inner.for.cond:
+// CHECK-NEXT: [[TMP13:%.*]] = load i32, ptr [[DOTOMP_IV]], align 4
+//
+// CHECK-LABEL: @main{{.*}}
Index: clang/lib/Sema/SemaOpenMP.cpp
===================================================================
--- clang/lib/Sema/SemaOpenMP.cpp
+++ clang/lib/Sema/SemaOpenMP.cpp
@@ -635,6 +635,11 @@
const SharingMapTy *Top = getTopOfStackOrNull();
return Top ? Top->Directive : OMPD_unknown;
}
+ void setCurrentDirective(OpenMPDirectiveKind newDK) {
+ SharingMapTy *Top = (SharingMapTy *)getTopOfStackOrNull();
+ assert(Top != NULL);
+ Top->Directive = newDK;
+ }
/// Returns directive kind at specified level.
OpenMPDirectiveKind getDirective(unsigned Level) const {
assert(!isStackEmpty() && "No directive at specified level.");
@@ -6120,10 +6125,75 @@
isOpenMPTargetDataManagementDirective(Kind)))
Diag(StartLoc, diag::warn_hip_omp_target_directives);
+ llvm::SmallVector<OMPClause *, 8> ClausesWithoutBind;
+ bool UseClausesWithoutBind = false;
+
+ // Restricting to "#pragma omp loop bind"
+ if (Kind == OMPD_loop) {
+ if (BindKind == OMPC_BIND_unknown) {
+ // Setting the enclosing teams or parallel construct for the loop
+ // directive without bind clause.
+ BindKind = OMPC_BIND_thread; // Default bind(thread) if binding is unknown
+
+ const OpenMPDirectiveKind ParentDirective =
+ DSAStack->getParentDirective();
+ if (ParentDirective == OMPD_unknown) {
+ Diag(DSAStack->getDefaultDSALocation(),
+ diag::err_omp_bind_required_on_loop);
+ } else if (ParentDirective == OMPD_parallel ||
+ ParentDirective == OMPD_target_parallel) {
+ BindKind = OMPC_BIND_parallel;
+ } else if (ParentDirective == OMPD_teams ||
+ ParentDirective == OMPD_target_teams) {
+ BindKind = OMPC_BIND_teams;
+ }
+ } else {
+ // bind clause is present, so we should set flag indicating to only
+ // use the clauses that aren't the bind clause for the new directive that
+ // loop is lowered to.
+ UseClausesWithoutBind = true;
+ }
+
+ for (OMPClause *C : Clauses) {
+ // Spec restriction : bind(teams) and reduction not permitted.
+ if (BindKind == OMPC_BIND_teams &&
+ C->getClauseKind() == llvm::omp::Clause::OMPC_reduction)
+ Diag(DSAStack->getDefaultDSALocation(),
+ diag::err_omp_loop_reduction_clause);
+
+ // A new Vector ClausesWithoutBind, which does not contain the bind
+ // clause, for passing to new directive.
+ if (C->getClauseKind() != llvm::omp::Clause::OMPC_bind)
+ ClausesWithoutBind.push_back(C);
+ }
+
+ switch (BindKind) {
+ case OMPC_BIND_parallel:
+ Kind = OMPD_for;
+ DSAStack->setCurrentDirective(OMPD_for);
+ break;
+ case OMPC_BIND_teams:
+ Kind = OMPD_distribute;
+ DSAStack->setCurrentDirective(OMPD_distribute);
+ break;
+ case OMPC_BIND_thread:
+ Kind = OMPD_simd;
+ DSAStack->setCurrentDirective(OMPD_simd);
+ break;
+ case OMPC_BIND_unknown:
+ break;
+ }
+ }
+
llvm::SmallVector<OMPClause *, 8> ClausesWithImplicit;
VarsWithInheritedDSAType VarsWithInheritedDSA;
bool ErrorFound = false;
- ClausesWithImplicit.append(Clauses.begin(), Clauses.end());
+ if (UseClausesWithoutBind) {
+ ClausesWithImplicit.append(ClausesWithoutBind.begin(),
+ ClausesWithoutBind.end());
+ } else {
+ ClausesWithImplicit.append(Clauses.begin(), Clauses.end());
+ }
if (AStmt && !CurContext->isDependentContext() && Kind != OMPD_atomic &&
Kind != OMPD_critical && Kind != OMPD_section && Kind != OMPD_master &&
Kind != OMPD_masked && !isOpenMPLoopTransformationDirective(Kind)) {
Index: clang/include/clang/Basic/DiagnosticSemaKinds.td
===================================================================
--- clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -9784,6 +9784,11 @@
def warn_loop_ctrl_binds_to_inner : Warning<
"'%0' is bound to current loop, GCC binds it to the enclosing loop">,
InGroup<GccCompat>;
+def err_omp_bind_required_on_loop : Error<
+ "expected 'bind' clause for loop construct without an enclosing OpenMP "
+ "construct">;
+def err_omp_loop_reduction_clause : Error<
+ "reduction clause not handled with '#pragma omp loop bind(teams)'">;
def warn_break_binds_to_switch : Warning<
"'break' is bound to loop, GCC binds it to switch">,
InGroup<GccCompat>;
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
