[Clang][OpenMP] Support for Code Generation of loop bind clause

[koops]

Summary

Support for Code Generation of "#pragma loop bind" clause.

1. bind(parallel)
2. bind(teams)
3. bind(thread)

[koops]

bind(thread) is not working at present. I have uploaded this patch to obtain feedback mainly on this.

[jdoerfert]

We need tests.

clang/lib/CodeGen/CGStmtOpenMP.cpp

7790 ↗

(On Diff #499780)

style, also below.

[ronlieb]

LIT Test ?

[ddpagan]

How is bind(thread) not working? Runtime issue?

[carlo.bertolli]

clang/lib/CodeGen/CGStmtOpenMP.cpp

7791 ↗

(On Diff #499780)

What if you have something like this:

void foo(..) {

#pragma omp loop
for (..) {
}

}

int main() {

#pragma omp target teams
{
  ...
  foo();
 }

#pragma omp target
 {
    #pragma omp parallel
    {
       foo();
    }

}

In the first invocation, loop is bound to teams. In the second, loop is bound to parallel.
This is a runtime condition.

I *believe* that not even OpenMP 6.0 TR1 allows us to decide at compile time if loop should be treated as worksharing or workdistribution....but I might be wrong.

Thanks!

[koops]

clang/lib/CodeGen/CGStmtOpenMP.cpp

7791 ↗

(On Diff #499780)

The loop directive does not have a bind clause. If you examine the current patch it takes care to preserve the old structure as is and the binding is done by default during runtime.

[carlo.bertolli]

clang/lib/CodeGen/CGStmtOpenMP.cpp

7791 ↗

(On Diff #499780)

Do you mean that this patch adds up to existing support - adding support for bind - and that existing support is already based on runtime support to decide on the case I mentioned?

[carlo.bertolli]

clang/lib/CodeGen/CGStmtOpenMP.cpp

7791 ↗

(On Diff #499780)

Disregard my comment: essentially, loop cannot be orphaned in the way I described.

[koops]

1. formatting
2. Adding lit test
3. Removing bind clause from the set of clauses passed during bind(parallel) to the OMPForDirective and bind(teams) to the OMPDistributeDirective.

[ABataev]

clang/lib/AST/StmtOpenMP.cpp

2362 ↗

(On Diff #505869)

We use tail allocation, this is definetely a mem leak.

clang/lib/CodeGen/CGStmtOpenMP.cpp

7815–7851 ↗

(On Diff #505869)

This is a bad place for this kind of operation here, it should be handled in Sema

[koops]

Moving the code to SemaOpenMP.cpp from CodeGen.

[koops]

Removing changes from :

1. clang/include/clang/AST/StmtOpenMP.h
2. clang/lib/AST/StmtOpenMP.cpp
3. clang/include/clang/Parse/Parser.h

These were useful when the code was in CodeGen to handle the bind clause.

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

6124

Remove this new line

6181–6189

1. You're overriding the directive kind here and do not restore it then. It may cause the compiler crash.
2. I think you need to here to create a new OpenMP region rather than overriding the existing one.

clang/test/OpenMP/loop_bind_codegen.cpp

1

Remove this

51–56

I think it should trigger the assert in setCurrentDirective

[koops]

clang/test/OpenMP/loop_bind_codegen.cpp

51–56

I did not see any crash at this point. Can you please give me other examples where I can see this crash?

[ABataev]

clang/include/clang/Basic/DiagnosticSemaKinds.td

9792

err_omp_...

clang/lib/Sema/SemaOpenMP.cpp

6146

ParentDirective

6167–6168

Remove extra parens

6173–6175

Remove extra braces

6181–6189

I see why you're doing this now. Would be good to see some semantic tests, like nesting of regions, use of unsupported clauses (e.g. linear on loop bind(teams)), etc.

6199–6202

Why need to drop bind clause here?

clang/test/OpenMP/loop_bind_codegen.cpp

51–56

Ah, you're overriding the loop directive kind.

[koops]

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

[ABataev]

In D144634#4272857, @koops wrote:

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

But why need to drop it? It makes processing more complex. The bind clause itself should not be a problem, no?

clang/lib/Sema/SemaOpenMP.cpp

638

NewDK

640

Add message to the assert.

clang/test/OpenMP/loop_bind_codegen.cpp

3–4

In other tests, this is used for testing with precompiled headers. You don't have this in your test runs currently.

9–14

Remove this commented code too

clang/test/OpenMP/loop_bind_messages.cpp

1–2 ↗

(On Diff #514096)

1. No need for this in semantic test

2, More tests required, nesting of regions, more tests for incompatible clauses.

[ABataev]

Also, add AST printing tests

[koops]

1. Taken care of some of the comments by Alexey.
2. Added extra tests of loop_bind_messages.cpp

However, generic_loop_messages.cpp & generic_loop_ast_print.cpp are present and provide a good coverage.

[ABataev]

In D144634#4308872, @koops wrote:

1. Taken care of some of the comments by Alexey.
2. Added extra tests of loop_bind_messages.cpp

However, generic_loop_messages.cpp & generic_loop_ast_print.cpp are present and provide a good coverage.

I rather doubt that these tests provide good coverage, since you're changing the directive kind here on the fly. This is a very new functionality, which was not tested before.
Add the tests for nesting of the regions, i.e. loop bind directive, enclosed in different regions, and diagnostics.

clang/lib/Sema/SemaOpenMP.cpp

339

What Ma1D stands for, what is the meaning of this abbrev? Add a comment.

clang/test/OpenMP/loop_bind_codegen.cpp

3

Add tests with save/load precompiled code.

clang/test/OpenMP/loop_bind_messages.cpp

3 ↗

(On Diff #518341)

Add tests with save/load precompiled code.

[koops]

However, generic_loop_messages.cpp & generic_loop_ast_print.cpp are present and provide a good coverage.

I rather doubt that these tests provide good coverage, since you're changing the directive kind here on the fly. This is a very new functionality, which was not tested before.
Add the tests for nesting of the regions, i.e. loop bind directive, enclosed in different regions, and diagnostics.

In https://reviews.llvm.org/D145823 : Add codegen for combined 'loop' directives, I see the following code
void CodeGenFunction::EmitOMPParallelGenericLoopDirective() {
....
....

emitCommonOMPParallelDirective(*this, S, OMPD_for, CodeGen, emitEmptyBoundParameters);

....
....
}
Here, the GenericLoopDirective is being changed to OMPD_for because of the Parallel Directive. Similarly,

void CodeGenFunction::EmitOMPTeamsGenericLoopDirective() {
....
....

CGF.CGM.getOpenMPRuntime().emitInlinedDirective(CGF, OMPD_distribute, CodeGenDistribute);

....
....
}
Here, GenericLoopDirective is being mapped to OMPD_distribute because of the Teams construct. These changes are very similar to mine. I still do not understand why you are saying "This is a very new functionality". If you can at least provide me some directions on which are all the "enclosed in different regions" it will be helpful.

[koops]

1. Addition of extra test case loop_bind_enclosed.cpp.
2. Changing of name from Map1D to MappedDirective

[koops]

This comment was removed by koops.

[ABataev]

clang/include/clang/Sema/Sema.h

11083 ↗

(On Diff #527127)

checkLastPrivateForMappedDirectives

clang/lib/AST/StmtPrinter.cpp

745–746 ↗

(On Diff #527127)

Just MappedDirective == llvm::omp::OMPD_loop

746–748 ↗

(On Diff #527127)

Remove braces

765–766 ↗

(On Diff #527127)

Just MappedDirective == llvm::omp::OMPD_loop

766–768 ↗

(On Diff #527127)

Remove braces

1005–1006 ↗

(On Diff #527127)

Just MappedDirective == llvm::omp::OMPD_loop

1006–1008 ↗

(On Diff #527127)

Remove braces

clang/lib/Sema/SemaOpenMP.cpp

342

I assume it shall be a member of SharingMapType, otherwise there might be problems with handling of enclosed constructs.

640

Why need a cast here?

641–642

use nullptr or just assert(Top &&

6136–6197

Could you outline this new code as a separate function?

10403

tmp is bad name. Give better name and follow LLVM coding style recommendations.

10438

Same

10449

Use c++ style of comments

10450–10451

if (DSAStack->getMappedDirective() == OMPD_loop && checkGenericLoopLastprivate(*this, Clauses, OMPD_loop, DSAStack)). Avoid structured complexity.

13962

Same, c++

13963–13964

Same, avoid complexity

13963–13967

`return DSAStack->getMappedDirective() == OMPD_loop && checkGenericLoopLastprivate(*this, Clauses, OMPD_loop, DSAStack);

[koops]

1. Taken care of Alexy's comments.
2. Reverting back changes to StmtPrinter.cpp because state of AST needs to be shown as is to the developer of clang when -ast-print/-ast-dump is used. This also meant some changes to the test cases to reflect the Mapped directives (from "omp loop" to "omp for", "omp distribute" or "omp simd").

[jdoerfert]

@ABataev wdyt?

[ddpagan]

clang/include/clang/Basic/DiagnosticSemaKinds.td

9793

Should be "allowed".

clang/lib/Sema/SemaOpenMP.cpp

6121

This function only applies to OMPD_loop. If *Kind != OMPD_loop then should just return UseClauseWithoutBind.

clang/test/OpenMP/generic_loop_ast_print.cpp

26 ↗

(On Diff #529532)

The AST that is printed should be what was originally specified in the source.

clang/test/OpenMP/loop_bind_codegen.cpp

2

Should auto-generate the the IR checking for this test using llvm/utils/update_cc_test_checks.py after verifying that what's generated is correct.

[ABataev]

clang/include/clang/Sema/Sema.h

11094 ↗

(On Diff #529532)

const member function?
Add a comment

11095 ↗

(On Diff #529532)

Add a comment

11095 ↗

(On Diff #529532)

SmallVectorImpl<OMPClause *> & instead of pointer to fully specified SmallVector

[koops]

clang/test/OpenMP/generic_loop_ast_print.cpp

26 ↗

(On Diff #529532)

Since I am mapping the loop directive to new directives (in the Sema itself) I am trying to print the state of AST as is to the developer of clang when -ast-print/-ast-dump is used. Without this correct printing of the state of the AST (new directives)) the developer may wonder why the code is not behaving as he/she specified and also in which phase the code gets changed.

[koops]

clang/include/clang/Sema/Sema.h

11094 ↗

(On Diff #529532)

"checkLastPrivateForMappedDirectives" is calling a non-const function "checkGenericLoopLastprivate". Since, this is a pre-existing function and is being called from other locations, I cannot make "checkLastPrivateForMappedDirectives" a const.

[koops]

1. Taking care of Alexy & David suggestions: a) Using update_cc_test_checks.py to generate CHECK statements. b) Change in error message from "handled" to "allowed". c) Adding comments for the bind clause. d) Mangled names of the functions in the CHECK statements are more generic with regular expressions.

[ABataev]

clang/include/clang/Basic/DiagnosticSemaKinds.td

9793

'reduction'

clang/lib/Sema/SemaOpenMP.cpp

6116

Why ClausesWithoutBind passing by pointer? Use llvm::SmallVectorImpl<OMPClause *> &ClausesWithoutBind instead

6118

Same, pass by reference

13982–13983

Restore formatting

[koops]

Addressing Alexey's comments.
The name matching in loop_bind_messages.cpp & generic_loop_codegen.cpp tests changed to take care of the failures.

[ABataev]

clang/include/clang/Basic/DiagnosticSemaKinds.td

9790

Also, 'loop'

clang/include/clang/Sema/Sema.h

11136 ↗

(On Diff #540373)

llvm::SmallVectorImpl<OMPClause *> &ClausesWithoutBind

[koops]

1. Removing the size for the template constructor, llvm::SmallVectorImpl<OMPClause *> &ClausesWithoutBind.
2. In the diagnostic message of err_omp_bind_required_on_loop, using single quotes for the name 'loop' construct.
3. Minor changes to the tests clang/test/OpenMP/loop_bind_codegen.cpp & clang/test/OpenMP/loop_bind_enclosed.cpp to use a generic pattern matching expression for the names & reducing some checks.

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

342

Not done

[koops]

Moving the variable MappedDirective into structure SharingMapTy. Enough comments have been put in place to explain.
Modified some of the loop_bind tests to cater to windows and debian platform.

[ABataev]

clang/include/clang/AST/StmtOpenMP.h

291 ↗

(On Diff #544428)

OpenMPDirectiveKind PrevMappedDirective = llvm::omp::OMPD_unknown;

308 ↗

(On Diff #544428)

Remove initialization of prevMappedDirective with the default value from the constructor

616 ↗

(On Diff #544428)

const member function

clang/include/clang/Sema/Sema.h

11476 ↗

(On Diff #544428)

All these changed functions are called with the default value of this new prevMappedDirective paramer. Remove.

clang/lib/Sema/SemaOpenMP.cpp

6365

Can you just use DSAStack->getMappedDirective() instead of prevMappedDirective?

10438

auto *ForDirective

[koops]

Fixing Alexey's suggestions.

[koops]

Correction for clang-format errors.

[ABataev]

clang/lib/Sema/TreeTransform.h

1652 ↗

(On Diff #544688)

PrevMappedDirective

[koops]

1. Converting prevMapped to PrevMappedDirective
2. Reducing the CHECK patterns in loop_bind_enclosed.cpp to the most essential ones to succeed matching for the IR generation on windows.

[koops]

1. clang-format error in clang/include/clang/AST/StmtOpenMP.h corrected.
2. Correcting clang/test/OpenMP/loop_bind_enclosed.cpp to use regular expressions for names of functions to accomodate names generated on Microsoft Windows.

[ABataev]

clang/include/clang/AST/StmtOpenMP.h

291 ↗

(On Diff #545728)

I don't see where this field is stored/loaded for PCH support. You need add a support for this in Serialization.

611–614 ↗

(On Diff #545728)

Better to make it a part of Create member function rather than having a separate function for this.

[koops]

1. clang/test/OpenMP/loop_bind_enclosed.cpp : Converting main and function names within it to use regular expressions to accomodate windows platform.

i) Instead of calling setMappedDirective() after the the creation of the Directive, made the MappedDirective a parameter of the Create() method.
ii) setMappedDirective() is called from the Directive::Create()
iii) The inheritance is : OMPExecutableDirective --> OMPLoopBasedDirective --> OMPLoopDirective --> OMPSimdDirective, from OMPSimdDirective constructor it involves a lot of changes to initialize PrevMappedDirective variable in OMPExecutableDirective starting from OMPSimdDirective, even if the variable is "protected". Hence we call setMappedDirective() from OMPSimdDirective::Create().

[koops]

clang/include/clang/AST/StmtOpenMP.h

291 ↗

(On Diff #545728)

Since the ASTReader and ASTWriter are used by the developer of clang, it is preferable to show the internal state of the compiler as is, e.g. if "#pragma omp loop bind(parallel)" is changed to "#pragma omp for" then the ASTWriter has to show it as "#pragma omp for". I can change it to "#pragma omp loop bind(parallel)" using the MappedDirective that is stored however, that would be misleading to the developer of clang.

[ABataev]

clang/include/clang/AST/StmtOpenMP.h

291 ↗

(On Diff #545728)

How it is related to PCH? Try to serialize/deserialize the template and then try to create a new instance. There will be an issue since there is no serialization support, when you try to create new instance in the Rebuild... function (mapLoopConstruct will use OMPD_unknown as Prev value instead of oringal previous value, since it is not serialized/deserialized)

[koops]

1. Support for PCH Serilization/deserilization for the PrevMapLoopConstruct variable in OMPExecutableDirective.
2. Modification of clang/test/PCH/pragma-loop.cpp to include "#pragma omp loop bind" with different parameters for bind clause.

[ABataev]

LG

[koops]

In clang/test/OpenMP/loop_bind_enclosed.cpp Generalizing the CHECK pattern for aarch64-linux, s390x-linux and ppc64le-linux.

[koops]

Making the CHECK pattern generic to match platforms tested after committing changes.

[koops]

Can someone please check for MacOS? Yesterday when this support was committed, the CHECK statements in tests loop_bind_codegen.cpp and loop_bind_enclosed.cpp had failed on Mac and I received a comment as follows:

When relanding, please remember to put a link to the review in the commit message.

I have made the CHECK statements generic enough to match those on Mac with CHECK for the functions that the test were expected.

[thakis]

In D144634#4573483, @koops wrote:

Can someone please check for MacOS? Yesterday when this support was committed, the CHECK statements in tests loop_bind_codegen.cpp and loop_bind_enclosed.cpp had failed on Mac

I patched this in and verified that check-clang now passes on macOS. Thanks!

and I received a comment as follows:

When relanding, please remember to put a link to the review in the commit message.

To do this, run git commit --amend and put Differential Revision: https://reviews.llvm.org/D144634 at the bottom of the commit message. git log has many, many examples of this :)

clang/include/clang/Basic/DiagnosticSemaKinds.td

 // Statements.
 def err_continue_not_in_loop : Error<
   "'continue' statement not in loop statement">;
 def err_break_not_in_loop_or_switch : Error<
   "'break' statement not in loop or switch statement">;
 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 error_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>;
 def err_default_not_in_switch : Error<
   "'default' statement not in switch statement">;
 def err_case_not_in_switch : Error<"'case' statement not in switch statement">;
 def warn_bool_switch_condition : Warning<
   "switch condition has boolean value">, InGroup<SwitchBool>;

clang/lib/Sema/SemaOpenMP.cpp

private:
   Expr *OMPPredefinedAllocators[OMPAllocateDeclAttr::OMPUserDefinedMemAlloc] = {
       nullptr};
   /// Vector of previously encountered target directives
   SmallVector<SourceLocation, 2> TargetLocations;
   SourceLocation AtomicLocation;
   /// Vector of declare variant construct traits.
   SmallVector<llvm::omp::TraitProperty, 8> ConstructTraits;
 
 public:
   explicit DSAStackTy(Sema &S) : SemaRef(S) {}
 
   /// Sets omp_allocator_handle_t type.
   void setOMPAllocatorHandleT(QualType Ty) { OMPAllocatorHandleT = Ty; }
   /// Gets omp_allocator_handle_t type.
   QualType getOMPAllocatorHandleT() const { return OMPAllocatorHandleT; }
   /// Sets omp_alloctrait_t type.
   void setOMPAlloctraitT(QualType Ty) { OMPAlloctraitT = Ty; }
   /// Gets omp_alloctrait_t type.
   QualType getOMPAlloctraitT() const { return OMPAlloctraitT; }
   /// Sets the given default allocator.
bool hasDirective(
           DPred,
       bool FromParent) const;
 
   /// Returns currently analyzed directive.
   OpenMPDirectiveKind getCurrentDirective() const {
     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.");
     return getStackElemAtLevel(Level).Directive;
   }
   /// Returns the capture region at the specified level.
   OpenMPDirectiveKind getCaptureRegion(unsigned Level,
                                        unsigned OpenMPCaptureLevel) const {
if (OMPClause *NewClause = S.ActOnOpenMPMapClause(
             MapperIdScopeSpec, MapperId, C->getMapType(),
             /*IsMapTypeImplicit=*/true, SourceLocation(), SourceLocation(),
             SubExprs, OMPVarListLocTy()))
       Clauses.push_back(NewClause);
   }
 }
 
 StmtResult Sema::ActOnOpenMPExecutableDirective(
     OpenMPDirectiveKind Kind, const DeclarationNameInfo &DirName,
     OpenMPDirectiveKind CancelRegion, ArrayRef<OMPClause *> Clauses,
     Stmt *AStmt, SourceLocation StartLoc, SourceLocation EndLoc) {
   StmtResult Res = StmtError();
   OpenMPBindClauseKind BindKind = OMPC_BIND_unknown;
   if (const OMPBindClause *BC =
           OMPExecutableDirective::getSingleClause<OMPBindClause>(Clauses))
     BindKind = BC->getBindKind();
+
   // First check CancelRegion which is then used in checkNestingOfRegions.
   if (checkCancelRegion(*this, Kind, CancelRegion, StartLoc) ||
       checkNestingOfRegions(*this, DSAStack, Kind, DirName, CancelRegion,
                             BindKind, StartLoc))
     return StmtError();
 
   // Report affected OpenMP target offloading behavior when in HIP lang-mode.
   if (getLangOpts().HIP && (isOpenMPTargetExecutionDirective(Kind) ||
                             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(SourceLocation(), diag::error_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;
+  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)) {
     assert(isa<CapturedStmt>(AStmt) && "Captured statement expected");
 
     // Check default data sharing attributes for referenced variables.
     DSAAttrChecker DSAChecker(DSAStack, *this, cast<CapturedStmt>(AStmt));
     int ThisCaptureLevel = getOpenMPCaptureLevels(Kind);
StmtResult Sema::ActOnOpenMPExecutableDirective(
   switch (Kind) {
   case OMPD_parallel:
     Res = ActOnOpenMPParallelDirective(ClausesWithImplicit, AStmt, StartLoc,
                                        EndLoc);
     AllowedNameModifiers.push_back(OMPD_parallel);
     break;
   case OMPD_simd:
     Res = ActOnOpenMPSimdDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc,
                                    VarsWithInheritedDSA);
     if (LangOpts.OpenMP >= 50)
       AllowedNameModifiers.push_back(OMPD_simd);
     break;
   case OMPD_tile:
     Res =
         ActOnOpenMPTileDirective(ClausesWithImplicit, AStmt, StartLoc, EndLoc);
     break;
   case OMPD_unroll:
if (!CurContext->isDependentContext()) {
     }
   }
 
   if (checkSimdlenSafelenSpecified(*this, Clauses))
     return StmtError();
 
   setFunctionHasBranchProtectedScope();
   return OMPSimdDirective::Create(Context, StartLoc, EndLoc, NestedLoopCount,
                                   Clauses, AStmt, B);
 }
 
 StmtResult
 Sema::ActOnOpenMPForDirective(ArrayRef<OMPClause *> Clauses, Stmt *AStmt,
                               SourceLocation StartLoc, SourceLocation EndLoc,
                               VarsWithInheritedDSAType &VarsWithImplicitDSA) {
   if (!AStmt)
     return StmtError();
for (OMPClause *C : Clauses) {
         if (FinishOpenMPLinearClause(*LC, cast<DeclRefExpr>(B.IterationVarRef),
                                      B.NumIterations, *this, CurScope,
                                      DSAStack))
           return StmtError();
     }
   }
 
   setFunctionHasBranchProtectedScope();
   return OMPForDirective::Create(
       Context, StartLoc, EndLoc, NestedLoopCount, Clauses, AStmt, B,
       DSAStack->getTaskgroupReductionRef(), DSAStack->isCancelRegion());
 }
 
 StmtResult Sema::ActOnOpenMPForSimdDirective(
     ArrayRef<OMPClause *> Clauses, Stmt *AStmt, SourceLocation StartLoc,
     SourceLocation EndLoc, VarsWithInheritedDSAType &VarsWithImplicitDSA) {
   if (!AStmt)
     return StmtError();
 
   assert(isa<CapturedStmt>(AStmt) && "Captured statement expected");
   OMPLoopBasedDirective::HelperExprs B;
   // In presence of clause 'collapse' or 'ordered' with number of loops, it will
   // define the nested loops number.
   unsigned NestedLoopCount =
       checkOpenMPLoop(OMPD_for_simd, getCollapseNumberExpr(Clauses),
                       getOrderedNumberExpr(Clauses), AStmt, *this, *DSAStack,
                       VarsWithImplicitDSA, B);
   if (NestedLoopCount == 0)
     return StmtError();
 
StmtResult Sema::ActOnOpenMPParallelMaskedTaskLoopSimdDirective(
 
   setFunctionHasBranchProtectedScope();
   return OMPParallelMaskedTaskLoopSimdDirective::Create(
       Context, StartLoc, EndLoc, NestedLoopCount, Clauses, AStmt, B);
 }
 
 StmtResult Sema::ActOnOpenMPDistributeDirective(
     ArrayRef<OMPClause *> Clauses, Stmt *AStmt, SourceLocation StartLoc,
     SourceLocation EndLoc, VarsWithInheritedDSAType &VarsWithImplicitDSA) {
   if (!AStmt)
     return StmtError();
 
   assert(isa<CapturedStmt>(AStmt) && "Captured statement expected");
   OMPLoopBasedDirective::HelperExprs B;
   // In presence of clause 'collapse' with number of loops, it will
   // define the nested loops number.
   unsigned NestedLoopCount =
       checkOpenMPLoop(OMPD_distribute, getCollapseNumberExpr(Clauses),
                       nullptr /*ordered not a clause on distribute*/, AStmt,
                       *this, *DSAStack, VarsWithImplicitDSA, B);
   if (NestedLoopCount == 0)
     return StmtError();
 
   assert((CurContext->isDependentContext() || B.builtAll()) &&
          "omp for loop exprs were not built");
 
   setFunctionHasBranchProtectedScope();
   return OMPDistributeDirective::Create(Context, StartLoc, EndLoc,
                                         NestedLoopCount, Clauses, AStmt, B);
 }
 
 StmtResult Sema::ActOnOpenMPDistributeParallelForDirective(
     ArrayRef<OMPClause *> Clauses, Stmt *AStmt, SourceLocation StartLoc,
     SourceLocation EndLoc, VarsWithInheritedDSAType &VarsWithImplicitDSA) {
   if (!AStmt)
     return StmtError();
 
   auto *CS = cast<CapturedStmt>(AStmt);

clang/test/OpenMP/loop_bind_codegen.cpp

+// Copyright 2020 Hewlett Packard Enterprise Development LP
+
+// 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 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_loop2() {
+  #pragma omp loop bind(thread)
+  for (int j = 0 ; j < NNN ; j++) {
+    aaa[j] = j*NNN;
+  }
+}
+
+int main() {
+  parallel_loop();
+  teams_loop();
+  thread_loop();
+  thread_loop2();
+
+  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-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 @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 {{[^@]+}}@_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.)
+// CHECK-LABEL: define {{[^@]+}}@.omp_outlined
+// CHECK:       omp.inner.for.cond:
+//
+// CHECK-LABEL: @main{{.*}}