Generate the capture for field when the field is used in openmp region with implicit default in the member function.

[jyu2]

Summary

This is to fix assert when field is referenced in openmp right with default (first|private) in member function.

The problem of assert is the capture is not generated for the field.

This patch is to generate capture when the field is used with implicit default, use it in the code, and save the capture off to make sure it is considered from that point and add first/private clauses.

1> Add new field ImplicitDefaultFirstprivateFDs in SharingMapTy, to store

generated capture fields info.

2> In function isOpenMPCaptureDecl: the caputer is generated and saved

in ImplicitDefaultFirstprivateFDs.

3> Add new help functions:

getImplicitFDCapExprDecl
isImplicitDefaultFirstprivateFD
addImplicitDefaultFirstprivateFD

4> Add addition argument in hasDSA to check default attribute for

default(first|private).

5> The isImplicitDefaultFirstprivateFD is used in VisitDeclRefExpr and

when building the implicit clause.

6> Add new parameter "Context" for buildCaptureDecl, due to when capture

field, the parent context is needed to be used.

7> Change isOpenMPPrivateDecl where stop propagate the capture from the

enclosing region for private variable.

8> In ActOnOpenMPFirstprivate/ActOnOpenMPPrivate, using captured info

to generate first|private clause.

[ABataev]

Why do we need to insert new implicit DSA? Only explicit DSAs are expected to be stored, implicit ones can be deduced using the rules.

[jyu2]

Thanks Alexey's comment. Address his comment. Instead save info into DSA, add new field in ImplicitDefaultFirstprivateFDs in SharingMapTy.

[ABataev]

clang/lib/Sema/SemaExprMember.cpp

1873–1877

Why do we need this function? Implicit private rule should apply (if should) only to this-Юашдув kind of expressions, if something like parent.field expression is used, parent should be private, not private.field. Or I'm missing something?

clang/lib/Sema/SemaOpenMP.cpp

201

///

1135–1136

1. Use /// style of comment.
2. const member function?

1147

const auto &?

1152–1153

1. Use /// style of comment.
2. const member function?

1161

const auto &?

1166

/// style

1173–1174

just emplace_back(FD, Sz, VD);?

2415

Better to add an actual param type here but without Param name

[jyu2]

Thanks Alexey for the review. This is address Alexey's comments.

[jyu2]

clang/lib/Sema/SemaExprMember.cpp

1873–1877

You are right for "parent.a" only privatize "parent".

But if 'a' is a field access in a member function then 'a' is privatized, not "this". It is same with firstprivate(a). But for the explicit firstprivate(a), the capture is build in ActOnOpenMPFirstprivateClause, so it can be mapped to reference in the omp region. For Implicit, I need to build capture for the first reference in the omp region with defalut(first|private) is used. And used that to generate firstprivate clause at end when the call to ActOnOpenMPFirstprivateClause when generating implicit clause.

Is this reasonable?

clang/lib/Sema/SemaOpenMP.cpp

201

Thanks Changed

2415

Changed.

[ABataev]

clang/lib/Sema/SemaExprMember.cpp

1873–1877

But you don't need all this info to build the capture. I think you can reuse existing isOpenMPCapturedDecl function without adding this extra isOpenMPFDCaptureDecl. Most probably, you don't need Base and all other stuff to build this->fd member expression.

[jyu2]

Thanks Alexey's review. Address Alexey's review.

clang/lib/Sema/SemaExprMember.cpp

1873–1877

Okay, I moved my code into isOpenMPCapturedDecl.

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

2446

Maybe BuildCXXThisExpr(SourceLocation(), getCurrentThisType(), /*IsImplicit=*/true)?

2458

Just DVarPrivate.CKind != OMPC_private

clang/lib/Sema/TreeTransform.h

11071–11074

Why do we need this check here?

[jyu2]

Address Alexey's comments! Thanks.

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

2446

Thanks. Change

2458

Yes!! changed

clang/lib/Sema/TreeTransform.h

11071–11074

Without, the field is not getting rebuild during the template instantiation. That cause the field is not getting captured and implicit firstprivate clause is not getting generate. The test is added in default_firstprivate_ast_print.cpp
where I add check for it on line 58:
DUMP-NEXT: -OMPFirstprivateClause
DUMP-NEXT: -DeclRefExpr {{.*}} 'targetDev'

[ABataev]

clang/lib/Sema/TreeTransform.h

11071–11074

This should work without changes. Do we need clause in the template function?

[jyu2]

clang/lib/Sema/TreeTransform.h

11071–11074

Do you mean, we don't need generate clause for instantiation's template function?

apply<32>
without this change the dump like:

| | `-CXXMethodDecl 0x1617ab8 <line:38:3, line:50:3> line:38:8 imported used apply 'void ()'
| |   |-TemplateArgument integral 32
| |   `-CompoundStmt 0x161c3a0 <col:16, line:50:3>
| |     `-OMPParallelDirective 0x161c350 <line:39:1, col:43>
| |       |-OMPDefaultClause 0x161bf30 <col:22, col:42>
| |       `-CapturedStmt 0x161c310 <line:40:5, line:42:5>

With this change
| | `-CXXMethodDecl 0x110cab8 <line:38:3, line:50:3> line:38:8 imported used apply 'void ()'
| |   |-TemplateArgument integral 32
| |   `-CompoundStmt 0x1111860 <col:16, line:50:3>
| |     `-OMPParallelDirective 0x1111808 <line:39:1, col:43>
| |       |-OMPDefaultClause 0x1110f30 <col:22, col:42>
| |       |-OMPFirstprivateClause 0x11117c8 <<invalid sloc>> <implicit>
| |       | `-DeclRefExpr 0x1111790 <line:41:7> 'int' lvalue OMPCapturedExpr 0x1111320 'targetDev' 'int &'
| |       `-CapturedStmt 0x1111590 <line:40:5, line:42:5>

[ABataev]

clang/lib/Sema/TreeTransform.h

11071–11074

It should work without this change correctly. If not, need to adjust sema analysis for such templated functions/fields.

[jyu2]

Remove change in TreeTransform.h. Will deal that in other patch. Thanks.

clang/lib/Sema/TreeTransform.h

11071–11074

Okay, I remove that change. Will deal that later.
It seems I need to set AlwaysRebuild to true. But I don't know how that works at this moment. Will submit other patch for this.

Thanks.

[ABataev]

clang/lib/Sema/TreeTransform.h

11071–11074

No, need to tweak the function that builds implicit clauses to build such clauses for non-templated decls, even in the template context

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

201

What is Sz here? Better to give a better name and add a description for the struct and all fields

1143

Is it possible to have an overflow here?

1148

What if Sz == -1?

1176

What about overflow here?

17493–17503

A check here not for curcontext dependent but for FD being dependent?

17772

Adjust this check for dependent context with non-dependent FD?

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

201

How about use StackLevel. Sorry use 0 instead

1143

I don't think that is possible. when I != EndI and Stacklevel should > 0 Or if I == EndI and Stacklevel should be 0

I add assert to check boundary condition be ensure:
assert((StackLevel > 0 && I != EndI) || (StackLevel == 0 && I == EndI));

1148

Should not be 0.

1176

Add assert for boundary check.

17493–17503

I can not this to work. Since the private copy only build under only build non-dependent context.

[jyu2]

Thanks Alexey for the review.
Add change in TreeTransform.h to rebuilt member expression.

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

199

Add comments for the structure and all fields, please.

2446

Do we need to check if getCurrentThisType() returns nullptr?

clang/lib/Sema/TreeTransform.h

11082

Can we also check if the field should be made private, i.e. avoid this short circuit if it is not in OpenMP region or is not supposed to be perivitized in the the context (no default private/firstprivate in the region)?

[jyu2]

Thanks Alexey for the review.
This is to address Alexey's comments,

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

199

Data required?

200

default

212

Captured?

clang/lib/Sema/TreeTransform.h

122

I think we don't need to add a new field here. Can instead we have a check for regions with default clauses, if possible?

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

17493–17503

fix typo:
I can not get this to work, since the private copy only build under non-dependent context.

clang/lib/Sema/TreeTransform.h

122

It seems during the TransformMemberExpr, I can not get Directives info for omp regions to check. Could you give me hand?
Thanks.

[ABataev]

clang/lib/Sema/TreeTransform.h

122

Yes, we do not rebuild the DSA stack at the time of the instantiation. Can you just check that we're inside OpenMPCapturedRegion? Something like getCurCapturedRegion()->CapRegionKind == CR_OpenMP? Or walk the stack of regions to find the outer OpenMP region, if any.

[jyu2]

Thanks Alexey for the review. This patch to address his comment.

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

212

Fixed. Thanks.

clang/lib/Sema/TreeTransform.h

122

Thank you so much for the help. I add new function in Sema to do this. Hope that is okay with you.

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

2265–2266

What if we have another outer OpenMP region, something like lambda inside OpenMP region?

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

2265–2266

My understanding is that hasDSA will go up to find innermost openmp region which has default clause. Am I right here?

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

2265–2266

Yes, if you're immediate captured region is OpenMP region. But what if you're inside lambda, which is inside OpenMP region? In this case getCurCapturedRegion()->CapRegionKind != CR_OpenMP. Will it still work, could add a test for this situation?

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

2265–2266

Good catch Thank you so much You are right. I removed check for CR_OpenMP. Just let hasDSA to find out. And add test for this.

[jyu2]

Thank Alexey for the review. This is remove check for omp region just use hasDSA.

[ABataev]

LG

[mgabka]

I noticed that this patch is causing now an assertion failure for cases like :

class A{

void a() {
  #pragma omp parallel
  a(b);
}

};

The failed assertion is: "const clang::ValueDecl* getCanonicalDecl(const clang::ValueDecl*): Assertion `FD' failed."

while before it clang was correctly reporting error:
error: use of undeclared identifier 'b'

is it the same assertion you were trying to fix here?

[jyu2]

In D127803#3693301, @mgabka wrote:

I noticed that this patch is causing now an assertion failure for cases like :

class A{

void a() {
  #pragma omp parallel
  a(b);
}

};

The failed assertion is: "const clang::ValueDecl* getCanonicalDecl(const clang::ValueDecl*): Assertion `FD' failed."

while before it clang was correctly reporting error:
error: use of undeclared identifier 'b'

is it the same assertion you were trying to fix here?

No, the assert I am fixing is when default(firstprivate) is used inside member function.

[ABataev]

In D127803#3693655, @jyu2 wrote:

In D127803#3693301, @mgabka wrote:

I noticed that this patch is causing now an assertion failure for cases like :

class A{

void a() {
  #pragma omp parallel
  a(b);
}

};

The failed assertion is: "const clang::ValueDecl* getCanonicalDecl(const clang::ValueDecl*): Assertion `FD' failed."

while before it clang was correctly reporting error:
error: use of undeclared identifier 'b'

is it the same assertion you were trying to fix here?

No, the assert I am fixing is when default(firstprivate) is used inside member function.

Could you please investigate and fix it?

[jyu2]

In D127803#3693660, @ABataev wrote:

In D127803#3693655, @jyu2 wrote:

In D127803#3693301, @mgabka wrote:

I noticed that this patch is causing now an assertion failure for cases like :

class A{

void a() {
  #pragma omp parallel
  a(b);
}

};

The failed assertion is: "const clang::ValueDecl* getCanonicalDecl(const clang::ValueDecl*): Assertion `FD' failed."

while before it clang was correctly reporting error:
error: use of undeclared identifier 'b'

is it the same assertion you were trying to fix here?

No, the assert I am fixing is when default(firstprivate) is used inside member function.

Could you please investigate and fix it?

Yes, I will fix this.

[mgabka]

In D127803#3693700, @jyu2 wrote:

In D127803#3693660, @ABataev wrote:

In D127803#3693655, @jyu2 wrote:

In D127803#3693301, @mgabka wrote:

I noticed that this patch is causing now an assertion failure for cases like :

class A{

void a() {
  #pragma omp parallel
  a(b);
}

};

The failed assertion is: "const clang::ValueDecl* getCanonicalDecl(const clang::ValueDecl*): Assertion `FD' failed."

while before it clang was correctly reporting error:
error: use of undeclared identifier 'b'

is it the same assertion you were trying to fix here?

No, the assert I am fixing is when default(firstprivate) is used inside member function.

Could you please investigate and fix it?

Yes, I will fix this.

Thanks a lot, I raised an upstream bug for it https://github.com/llvm/llvm-project/issues/56884

clang/include/clang/Sema/Sema.h

public:
   bool isOpenMPCapturedByRef(const ValueDecl *D, unsigned Level,
                              unsigned OpenMPCaptureLevel) const;
 
   /// Check if the specified variable is used in one of the private
   /// clauses (private, firstprivate, lastprivate, reduction etc.) in OpenMP
   /// constructs.
   VarDecl *isOpenMPCapturedDecl(ValueDecl *D, bool CheckScopeInfo = false,
                                 unsigned StopAt = 0);
+
+  VarDecl *isOpenMPFDCaptureDecl(ValueDecl *D, Expr *Base, bool IsArrow,
+                                 SourceLocation OpLoc, const CXXScopeSpec *SS,
+                                 SourceLocation TemplateKWLoc,
+                                 ValueDecl *Member, DeclAccessPair FoundDecl,
+                                 bool HadMultipleCandidates,
+                                 const DeclarationNameInfo &MemberNameInfo,
+                                 QualType Ty, ExprValueKind VK,
+                                 ExprObjectKind OK);
+
   ExprResult getOpenMPCapturedExpr(VarDecl *Capture, ExprValueKind VK,
                                    ExprObjectKind OK, SourceLocation Loc);
 
   /// If the current region is a loop-based region, mark the start of the loop
   /// construct.
   void startOpenMPLoop();
 
   /// If the current region is a range loop-based region, mark the start of the

clang/lib/Sema/SemaExprMember.cpp

Sema::BuildFieldReferenceExpr(Expr *BaseExpr, bool IsArrow,
   // non-static member functions, privatized by OpenMP constructs.
   if (getLangOpts().OpenMP && IsArrow &&
       !CurContext->isDependentContext() &&
       isa<CXXThisExpr>(Base.get()->IgnoreParenImpCasts())) {
     if (auto *PrivateCopy = isOpenMPCapturedDecl(Field)) {
       return getOpenMPCapturedExpr(PrivateCopy, VK, OK,
                                    MemberNameInfo.getLoc());
     }
+    if (auto *PrivateCopy =
+            isOpenMPFDCaptureDecl(Field, Base.get(), IsArrow, OpLoc, &SS,
+                                  /*TemplateKWLoc=*/SourceLocation(), Field,
+                                  FoundDecl, /*HadMultipleCandidates=*/false,
+                                  MemberNameInfo, MemberType, VK, OK))
+      return getOpenMPCapturedExpr(PrivateCopy, VK, OK,
+                                   MemberNameInfo.getLoc());
   }
 
   return BuildMemberExpr(Base.get(), IsArrow, OpLoc, &SS,
                          /*TemplateKWLoc=*/SourceLocation(), Field, FoundDecl,
                          /*HadMultipleCandidates=*/false, MemberNameInfo,
                          MemberType, VK, OK);
 }
 

clang/lib/Sema/SemaOpenMP.cpp

struct SharingMapTy {
     llvm::DenseSet<CanonicalDeclPtr<Decl>> ImplicitTaskFirstprivates;
     /// List of globals marked as declare target link in this target region
     /// (isOpenMPTargetExecutionDirective(Directive) == true).
     llvm::SmallVector<DeclRefExpr *, 4> DeclareTargetLinkVarDecls;
     /// List of decls used in inclusive/exclusive clauses of the scan directive.
     llvm::DenseSet<CanonicalDeclPtr<Decl>> UsedInScanDirective;
     llvm::DenseMap<CanonicalDeclPtr<const Decl>, UsesAllocatorsDeclKind>
         UsesAllocatorsDecls;
+    struct ImplicitDefaultFDInfoTy {
+      const FieldDecl *FD = nullptr;
+      size_t Sz = -1;
+      VarDecl *VD = nullptr;
+      ImplicitDefaultFDInfoTy(const FieldDecl *FD, size_t Sz, VarDecl *VD)
+          : FD(FD), Sz(Sz), VD(VD) {}
+    };
+    /// List of captuer fields
+    llvm::SmallVector<ImplicitDefaultFDInfoTy, 8>
+        ImplicitDefaultFirstprivateFDs;
     Expr *DeclareMapperVar = nullptr;
     SharingMapTy(OpenMPDirectiveKind DKind, DeclarationNameInfo Name,
                  Scope *CurScope, SourceLocation Loc)
         : Directive(DKind), DirectiveName(Name), CurScope(CurScope),
           ConstructLoc(Loc) {}
     SharingMapTy() = default;
   };
 
   using StackTy = SmallVector<SharingMapTy, 4>;
 
   /// Stack of used declaration and their data-sharing attributes.
   DeclSAMapTy Threadprivates;
public:
   const DSAVarData getImplicitDSA(ValueDecl *D, bool FromParent) const;
   /// Returns data-sharing attributes for the specified declaration.
   const DSAVarData getImplicitDSA(ValueDecl *D, unsigned Level) const;
   /// Checks if the specified variables has data-sharing attributes which
   /// match specified \a CPred predicate in any directive which matches \a DPred
   /// predicate.
   const DSAVarData
   hasDSA(ValueDecl *D,
-         const llvm::function_ref<bool(OpenMPClauseKind, bool)> CPred,
+         const llvm::function_ref<bool(OpenMPClauseKind, bool,
+                                       DefaultDataSharingAttributes)>
+             CPred,
          const llvm::function_ref<bool(OpenMPDirectiveKind)> DPred,
          bool FromParent) const;
   /// Checks if the specified variables has data-sharing attributes which
   /// match specified \a CPred predicate in any innermost directive which
   /// matches \a DPred predicate.
   const DSAVarData
   hasInnermostDSA(ValueDecl *D,
                   const llvm::function_ref<bool(OpenMPClauseKind, bool)> CPred,
public:
   void addDeclareMapperVarRef(Expr *Ref) {
     SharingMapTy &StackElem = getTopOfStack();
     StackElem.DeclareMapperVar = Ref;
   }
   const Expr *getDeclareMapperVarRef() const {
     const SharingMapTy *Top = getTopOfStackOrNull();
     return Top ? Top->DeclareMapperVar : nullptr;
   }
+  /// get captured field from ImplicitDefaultFirstprivateFDs
+  VarDecl *getImplicitFDCapExprDecl(const FieldDecl *FD) const {
+    const_iterator I = begin();
+    const_iterator EndI = end();
+    size_t Sz = getStackSize();
+    for (; I != EndI; ++I) {
+      if (I->DefaultAttr == DSA_firstprivate || I->DefaultAttr == DSA_private)
+        break;
+      Sz--;
+    }
+    if (I == EndI)
+      return nullptr;
+    for (const auto &IFD : I->ImplicitDefaultFirstprivateFDs)
+      if (IFD.FD == FD && IFD.Sz == Sz)
+        return IFD.VD;
+    return nullptr;
+  }
+  /// Check if capture decl is field captured in ImplicitDefaultFirstprivateFDs
+  bool isImplicitDefaultFirstprivateFD(VarDecl *VD) const {
+    const_iterator I = begin();
+    const_iterator EndI = end();
+    for (; I != EndI; ++I)
+      if (I->DefaultAttr == DSA_firstprivate || I->DefaultAttr == DSA_private)
+        break;
+    if (I == EndI)
+      return false;
+    for (const auto &IFD : I->ImplicitDefaultFirstprivateFDs)
+      if (IFD.VD == VD)
+        return true;
+    return false;
+  }
+  /// Store capture FD info in ImplicitDefaultFirstprivateFDs
+  void addImplicitDefaultFirstprivateFD(const FieldDecl *FD, VarDecl *VD) {
+    iterator I = begin();
+    const_iterator EndI = end();
+    size_t Sz = getStackSize();
+    for (; I != EndI; ++I) {
+      if (I->DefaultAttr == DSA_private || I->DefaultAttr == DSA_firstprivate) {
+        I->ImplicitDefaultFirstprivateFDs.emplace_back(FD, Sz, VD);
+        break;
+      }
+      Sz--;
+    }
+  }
 };
 
 bool isImplicitTaskingRegion(OpenMPDirectiveKind DKind) {
   return isOpenMPParallelDirective(DKind) || isOpenMPTeamsDirective(DKind);
 }
 
 bool isImplicitOrExplicitTaskingRegion(OpenMPDirectiveKind DKind) {
   return isImplicitTaskingRegion(DKind) || isOpenMPTaskingDirective(DKind) ||
if (getStackSize() <= Level)
     return DSAVarData();
   D = getCanonicalDecl(D);
   const_iterator StartI = std::next(begin(), getStackSize() - 1 - Level);
   return getDSA(StartI, D);
 }
 
 const DSAStackTy::DSAVarData
 DSAStackTy::hasDSA(ValueDecl *D,
-                   const llvm::function_ref<bool(OpenMPClauseKind, bool)> CPred,
+                   const llvm::function_ref<bool(OpenMPClauseKind, bool,
+                                                 DefaultDataSharingAttributes)>
+                       CPred,
                    const llvm::function_ref<bool(OpenMPDirectiveKind)> DPred,
                    bool FromParent) const {
   if (isStackEmpty())
     return {};
   D = getCanonicalDecl(D);
   const_iterator I = begin();
   const_iterator EndI = end();
   if (FromParent && I != EndI)
     ++I;
   for (; I != EndI; ++I) {
     if (!DPred(I->Directive) &&
         !isImplicitOrExplicitTaskingRegion(I->Directive))
       continue;
     const_iterator NewI = I;
     DSAVarData DVar = getDSA(NewI, D);
-    if (I == NewI && CPred(DVar.CKind, DVar.AppliedToPointee))
+    if (I == NewI && CPred(DVar.CKind, DVar.AppliedToPointee, I->DefaultAttr))
       return DVar;
   }
   return {};
 }
 
 const DSAStackTy::DSAVarData DSAStackTy::hasInnermostDSA(
     ValueDecl *D, const llvm::function_ref<bool(OpenMPClauseKind, bool)> CPred,
     const llvm::function_ref<bool(OpenMPDirectiveKind)> DPred,
return (isOpenMPTargetExecutionDirective(DSAStack->getCurrentDirective()) &&
          DSAStack->hasDirective(
              [](OpenMPDirectiveKind K, const DeclarationNameInfo &,
                 SourceLocation) -> bool {
                return isOpenMPTargetExecutionDirective(K);
              },
              false);
 }
 
+static OMPCapturedExprDecl *buildCaptureDecl(Sema &S, IdentifierInfo *Id,
+                                             Expr *CaptureExpr, bool WithInit,
+                                             DeclContext *CurContext,
+                                             bool AsExpression);
+
+VarDecl *Sema::isOpenMPFDCaptureDecl(
+    ValueDecl *D, Expr *Base, bool IsArrow, SourceLocation OpLoc,
+    const CXXScopeSpec *SS, SourceLocation TemplateKWLoc, ValueDecl *Member,
+    DeclAccessPair FoundDecl, bool HadMultipleCandidates,
+    const DeclarationNameInfo &MemberNameInfo, QualType Ty, ExprValueKind VK,
+    ExprObjectKind OK) {
+  assert(LangOpts.OpenMP && "OpenMP is not allowed");
+  D = getCanonicalDecl(D);
+
+  auto *FD = dyn_cast<FieldDecl>(D);
+  if (DSAStack->getCurrentDirective() != OMPD_unknown &&
+      (!DSAStack->isClauseParsingMode() ||
+       DSAStack->getParentDirective() != OMPD_unknown)) {
+    DSAStackTy::DSAVarData DVarPrivate = DSAStack->hasDSA(
+        D,
+        [](OpenMPClauseKind C, bool AppliedToPointee,
+           DefaultDataSharingAttributes DefaultAttr) {
+          return isOpenMPPrivate(C) && !AppliedToPointee &&
+                 (DefaultAttr == DSA_firstprivate ||
+                  DefaultAttr == DSA_private);
+        },
+        [](OpenMPDirectiveKind) { return true; },
+        DSAStack->isClauseParsingMode());
+    if (DVarPrivate.CKind == OMPC_unknown)
+      return nullptr;
+
+    VarDecl *VD = DSAStack->getImplicitFDCapExprDecl(FD);
+    if (!VD) {
+      ExprResult ME = BuildMemberExpr(Base, IsArrow, OpLoc, SS, TemplateKWLoc,
+                                      Member, FoundDecl, HadMultipleCandidates,
+                                      MemberNameInfo, Ty, VK, OK);
+      OMPCapturedExprDecl *CD =
+          buildCaptureDecl(*this, FD->getIdentifier(), ME.get(),
+                           DVarPrivate.CKind == OMPC_private ? false : true,
+                           CurContext->getParent(),
+                           /*AsExpression=*/false);
+      DeclRefExpr *VDPrivateRefExpr =
+          buildDeclRefExpr(*this, CD, CD->getType().getNonReferenceType(),
+                           MemberNameInfo.getLoc());
+      VD = cast<VarDecl>(VDPrivateRefExpr->getDecl());
+      DSAStack->addImplicitDefaultFirstprivateFD(FD, VD);
+      return VD;
+    }
+    return VD;
+  }
+  return nullptr;
+}
+
 VarDecl *Sema::isOpenMPCapturedDecl(ValueDecl *D, bool CheckScopeInfo,
                                     unsigned StopAt) {
   assert(LangOpts.OpenMP && "OpenMP is not allowed");
   D = getCanonicalDecl(D);
 
   auto *VD = dyn_cast<VarDecl>(D);
   // Do not capture constexpr variables.
   if (VD && VD->isConstexpr())
if (DVarTop.CKind != OMPC_unknown && isOpenMPPrivate(DVarTop.CKind) &&
       return VD ? VD : cast<VarDecl>(DVarTop.PrivateCopy->getDecl());
     // Threadprivate variables must not be captured.
     if (isOpenMPThreadPrivate(DVarTop.CKind))
       return nullptr;
     // The variable is not private or it is the variable in the directive with
     // default(none) clause and not used in any clause.
     DSAStackTy::DSAVarData DVarPrivate = DSAStack->hasDSA(
         D,
-        [](OpenMPClauseKind C, bool AppliedToPointee) {
+        [](OpenMPClauseKind C, bool AppliedToPointee, bool) {
           return isOpenMPPrivate(C) && !AppliedToPointee;
         },
         [](OpenMPDirectiveKind) { return true; },
         DSAStack->isClauseParsingMode());
     // Global shared must not be captured.
     if (VD && !VD->hasLocalStorage() && DVarPrivate.CKind == OMPC_unknown &&
         ((DSAStack->getDefaultDSA() != DSA_none &&
           DSAStack->getDefaultDSA() != DSA_private &&
           DSAStack->getDefaultDSA() != DSA_firstprivate) ||
          DVarTop.CKind == OMPC_shared))
       return nullptr;
     if (DVarPrivate.CKind != OMPC_unknown ||
         (VD && (DSAStack->getDefaultDSA() == DSA_none ||
                 DSAStack->getDefaultDSA() == DSA_private ||
                 DSAStack->getDefaultDSA() == DSA_firstprivate)))
-      return VD ? VD : cast<VarDecl>(DVarPrivate.PrivateCopy->getDecl());
+      return VD ? VD
+             : DVarPrivate.PrivateCopy
+                 ? cast<VarDecl>(DVarPrivate.PrivateCopy->getDecl())
+                 : nullptr;
   }
   return nullptr;
 }
 
 void Sema::adjustOpenMPTargetScopeIndex(unsigned &FunctionScopesIndex,
                                         unsigned Level) const {
   FunctionScopesIndex -= getOpenMPCaptureLevels(DSAStack->getDirective(Level));
 }
 
 void Sema::startOpenMPLoop() {
   assert(LangOpts.OpenMP && "OpenMP must be enabled.");
   if (isOpenMPLoopDirective(DSAStack->getCurrentDirective()))
     DSAStack->loopInit();
 }
 
 void Sema::startOpenMPCXXRangeFor() {
   assert(LangOpts.OpenMP && "OpenMP must be enabled.");
   if (isOpenMPLoopDirective(DSAStack->getCurrentDirective())) {
     DSAStack->resetPossibleLoopCounter();
     DSAStack->loopStart();
   }
 }
 
 OpenMPClauseKind Sema::isOpenMPPrivateDecl(ValueDecl *D, unsigned Level,
                                            unsigned CapLevel) const {
   assert(LangOpts.OpenMP && "OpenMP is not allowed");
+  if (DSAStack->getCurrentDirective() != OMPD_unknown &&
+      (!DSAStack->isClauseParsingMode() ||
+       DSAStack->getParentDirective() != OMPD_unknown)) {
+    DSAStackTy::DSAVarData DVarPrivate = DSAStack->hasDSA(
+        D,
+        [](OpenMPClauseKind C, bool AppliedToPointee,
+           DefaultDataSharingAttributes DefaultAttr) {
+          return isOpenMPPrivate(C) && !AppliedToPointee &&
+                 DefaultAttr == DSA_private;
+        },
+        [](OpenMPDirectiveKind) { return true; },
+        DSAStack->isClauseParsingMode());
+    if (DVarPrivate.CKind == OMPC_private && isa<OMPCapturedExprDecl>(D) &&
+        DSAStack->isImplicitDefaultFirstprivateFD(cast<VarDecl>(D)) &&
+        !DSAStack->isLoopControlVariable(D).first)
+      return OMPC_private;
+  }
   if (DSAStack->hasExplicitDirective(isOpenMPTaskingDirective, Level)) {
     bool IsTriviallyCopyable =
         D->getType().getNonReferenceType().isTriviallyCopyableType(Context) &&
         !D->getType()
              .getNonReferenceType()
              .getCanonicalType()
              ->getAsCXXRecordDecl();
     OpenMPDirectiveKind DKind = DSAStack->getDirective(Level);
public:
   void VisitDeclRefExpr(DeclRefExpr *E) {
     if (TryCaptureCXXThisMembers || E->isTypeDependent() ||
         E->isValueDependent() || E->containsUnexpandedParameterPack() ||
         E->isInstantiationDependent())
       return;
     if (auto *VD = dyn_cast<VarDecl>(E->getDecl())) {
       // Check the datasharing rules for the expressions in the clauses.
       if (!CS || (isa<OMPCapturedExprDecl>(VD) && !CS->capturesVariable(VD) &&
-                  !Stack->getTopDSA(VD, /*FromParent=*/false).RefExpr)) {
+                  !Stack->getTopDSA(VD, /*FromParent=*/false).RefExpr &&
+                  !Stack->isImplicitDefaultFirstprivateFD(VD))) {
         if (auto *CED = dyn_cast<OMPCapturedExprDecl>(VD))
           if (!CED->hasAttr<OMPCaptureNoInitAttr>()) {
             Visit(CED->getInit());
             return;
           }
       } else if (VD->isImplicit() || isa<OMPCapturedExprDecl>(VD))
         // Do not analyze internal variables and do not enclose them into
         // implicit clauses.
+        if (!Stack->isImplicitDefaultFirstprivateFD(VD))
           return;
       VD = VD->getCanonicalDecl();
       // Skip internally declared variables.
       if (VD->hasLocalStorage() && CS && !CS->capturesVariable(VD) &&
+          !Stack->isImplicitDefaultFirstprivateFD(VD) &&
           !Stack->isImplicitTaskFirstprivate(VD))
         return;
       // Skip allocators in uses_allocators clauses.
       if (Stack->isUsesAllocatorsDecl(VD))
         return;
 
       DSAStackTy::DSAVarData DVar = Stack->getTopDSA(VD, /*FromParent=*/false);
       // Check if the variable has explicit DSA set and stop analysis if it so.

 int Sema::getOpenMPCaptureLevels(OpenMPDirectiveKind DKind) {
   SmallVector<OpenMPDirectiveKind, 4> CaptureRegions;
   getOpenMPCaptureRegions(CaptureRegions, DKind);
   return CaptureRegions.size();
 }
 
 static OMPCapturedExprDecl *buildCaptureDecl(Sema &S, IdentifierInfo *Id,
                                              Expr *CaptureExpr, bool WithInit,
+                                             DeclContext *CurContext,
                                              bool AsExpression) {
   assert(CaptureExpr);
   ASTContext &C = S.getASTContext();
   Expr *Init = AsExpression ? CaptureExpr : CaptureExpr->IgnoreImpCasts();
   QualType Ty = Init->getType();
   if (CaptureExpr->getObjectKind() == OK_Ordinary && CaptureExpr->isGLValue()) {
     if (S.getLangOpts().CPlusPlus) {
       Ty = C.getLValueReferenceType(Ty);
     } else {
       Ty = C.getPointerType(Ty);
       ExprResult Res =
           S.CreateBuiltinUnaryOp(CaptureExpr->getExprLoc(), UO_AddrOf, Init);
       if (!Res.isUsable())
         return nullptr;
       Init = Res.get();
     }
     WithInit = true;
   }
-  auto *CED = OMPCapturedExprDecl::Create(C, S.CurContext, Id, Ty,
+  auto *CED = OMPCapturedExprDecl::Create(C, CurContext, Id, Ty,
                                           CaptureExpr->getBeginLoc());
   if (!WithInit)
     CED->addAttr(OMPCaptureNoInitAttr::CreateImplicit(C));
-  S.CurContext->addHiddenDecl(CED);
+  CurContext->addHiddenDecl(CED);
   Sema::TentativeAnalysisScope Trap(S);
   S.AddInitializerToDecl(CED, Init, /*DirectInit=*/false);
   return CED;
 }
 
 static DeclRefExpr *buildCapture(Sema &S, ValueDecl *D, Expr *CaptureExpr,
                                  bool WithInit) {
   OMPCapturedExprDecl *CD;
   if (VarDecl *VD = S.isOpenMPCapturedDecl(D))
     CD = cast<OMPCapturedExprDecl>(VD);
   else
     CD = buildCaptureDecl(S, D->getIdentifier(), CaptureExpr, WithInit,
+                          S.CurContext,
                           /*AsExpression=*/false);
   return buildDeclRefExpr(S, CD, CD->getType().getNonReferenceType(),
                           CaptureExpr->getExprLoc());
 }
 
 static ExprResult buildCapture(Sema &S, Expr *CaptureExpr, DeclRefExpr *&Ref) {
   CaptureExpr = S.DefaultLvalueConversion(CaptureExpr).get();
   if (!Ref) {
     OMPCapturedExprDecl *CD = buildCaptureDecl(
         S, &S.getASTContext().Idents.get(".capture_expr."), CaptureExpr,
-        /*WithInit=*/true, /*AsExpression=*/true);
+        /*WithInit=*/true, S.CurContext, /*AsExpression=*/true);
     Ref = buildDeclRefExpr(S, CD, CD->getType().getNonReferenceType(),
                            CaptureExpr->getExprLoc());
   }
   ExprResult Res = Ref;
   if (!S.getLangOpts().CPlusPlus &&
       CaptureExpr->getObjectKind() == OK_Ordinary && CaptureExpr->isGLValue() &&
       Ref->getType()->isPointerType()) {
     Res = S.CreateBuiltinUnaryOp(CaptureExpr->getExprLoc(), UO_Deref, Ref);

 }
 
 OMPClause *Sema::ActOnOpenMPPrivateClause(ArrayRef<Expr *> VarList,
                                           SourceLocation StartLoc,
                                           SourceLocation LParenLoc,
                                           SourceLocation EndLoc) {
   SmallVector<Expr *, 8> Vars;
   SmallVector<Expr *, 8> PrivateCopies;
+  bool IsImplicitClause =
+      StartLoc.isInvalid() && LParenLoc.isInvalid() && EndLoc.isInvalid();
   for (Expr *RefExpr : VarList) {
     assert(RefExpr && "NULL expr in OpenMP private clause.");
     SourceLocation ELoc;
     SourceRange ERange;
     Expr *SimpleRefExpr = RefExpr;
     auto Res = getPrivateItem(*this, SimpleRefExpr, ELoc, ERange);
     if (Res.second) {
       // It will be analyzed later.
VarDecl *VDPrivate =
                      VD ? cast<DeclRefExpr>(SimpleRefExpr) : nullptr);
     ActOnUninitializedDecl(VDPrivate);
     if (VDPrivate->isInvalidDecl())
       continue;
     DeclRefExpr *VDPrivateRefExpr = buildDeclRefExpr(
         *this, VDPrivate, RefExpr->getType().getUnqualifiedType(), ELoc);
 
     DeclRefExpr *Ref = nullptr;
-    if (!VD && !CurContext->isDependentContext())
+    if (!VD && !CurContext->isDependentContext()) {
+      auto *FD = dyn_cast<FieldDecl>(D);
+      VarDecl *VD = FD ? DSAStack->getImplicitFDCapExprDecl(FD) : nullptr;
+      if (VD)
+        Ref = buildDeclRefExpr(*this, VD, VD->getType().getNonReferenceType(),
+                               RefExpr->getExprLoc());
+      else
         Ref = buildCapture(*this, D, SimpleRefExpr, /*WithInit=*/false);
+    }
+    if (!IsImplicitClause)
       DSAStack->addDSA(D, RefExpr->IgnoreParens(), OMPC_private, Ref);
     Vars.push_back((VD || CurContext->isDependentContext())
                        ? RefExpr->IgnoreParens()
                        : Ref);
     PrivateCopies.push_back(VDPrivateRefExpr);
   }
 
   if (Vars.empty())
     return nullptr;
if (VDPrivate->isInvalidDecl()) {
       }
       continue;
     }
     CurContext->addDecl(VDPrivate);
     DeclRefExpr *VDPrivateRefExpr = buildDeclRefExpr(
         *this, VDPrivate, RefExpr->getType().getUnqualifiedType(),
         RefExpr->getExprLoc());
     DeclRefExpr *Ref = nullptr;
     if (!VD && !CurContext->isDependentContext()) {
       if (TopDVar.CKind == OMPC_lastprivate) {
         Ref = TopDVar.PrivateCopy;
       } else {
+        auto *FD = dyn_cast<FieldDecl>(D);
+        VarDecl *VD = FD ? DSAStack->getImplicitFDCapExprDecl(FD) : nullptr;
+        if (VD)
+          Ref = buildDeclRefExpr(*this, VD, VD->getType().getNonReferenceType(),
+                                 RefExpr->getExprLoc());
+        else
           Ref = buildCapture(*this, D, SimpleRefExpr, /*WithInit=*/true);
         if (!isOpenMPCapturedDecl(D))
           ExprCaptures.push_back(Ref->getDecl());
       }
     }
     if (!IsImplicitClause)
       DSAStack->addDSA(D, RefExpr->IgnoreParens(), OMPC_firstprivate, Ref);
     Vars.push_back((VD || CurContext->isDependentContext())
                        ? RefExpr->IgnoreParens()
bool VisitDeclRefExpr(DeclRefExpr *E) {
     if (auto *VD = dyn_cast<VarDecl>(E->getDecl())) {
       DSAStackTy::DSAVarData DVar = Stack->getTopDSA(VD, /*FromParent=*/false);
       if (DVar.CKind == OMPC_shared && !DVar.RefExpr)
         return false;
       if (DVar.CKind != OMPC_unknown)
         return true;
       DSAStackTy::DSAVarData DVarPrivate = Stack->hasDSA(
           VD,
-          [](OpenMPClauseKind C, bool AppliedToPointee) {
+          [](OpenMPClauseKind C, bool AppliedToPointee, bool) {
             return isOpenMPPrivate(C) && !AppliedToPointee;
           },
           [](OpenMPDirectiveKind) { return true; },
           /*FromParent=*/true);
       return DVarPrivate.CKind != OMPC_unknown;
     }
     return false;
   }

clang/lib/Sema/TreeTransform.h

public:
     ~ForgetPartiallySubstitutedPackRAII() {
       Self.RememberPartiallySubstitutedPack(Old);
     }
   };
 
 protected:
   Sema &SemaRef;
 
   /// The set of local declarations that have been transformed, for
   /// cases where we are forced to build new declarations within the transformer
   /// rather than in the subclass (e.g., lambda closure types).
   llvm::DenseMap<Decl *, Decl *> TransformedLocalDecls;
 
 public:
   /// Initializes a new tree transformer.
   TreeTransform(Sema &SemaRef) : SemaRef(SemaRef) { }
 
if (FoundDecl == E->getMemberDecl()) {
     FoundDecl = Member;
   } else {
     FoundDecl = cast_or_null<NamedDecl>(
                    getDerived().TransformDecl(E->getMemberLoc(), FoundDecl));
     if (!FoundDecl)
       return ExprError();
   }
 
-  if (!getDerived().AlwaysRebuild() &&
-      Base.get() == E->getBase() &&
-      QualifierLoc == E->getQualifierLoc() &&
-      Member == E->getMemberDecl() &&
-      FoundDecl == E->getFoundDecl() &&
+  if (!getDerived().AlwaysRebuild() && !getSema().getLangOpts().OpenMP &&
+      Base.get() == E->getBase() && QualifierLoc == E->getQualifierLoc() &&
+      Member == E->getMemberDecl() && FoundDecl == E->getFoundDecl() &&
       !E->hasExplicitTemplateArgs()) {
 
     // Mark it referenced in the new context regardless.
     // FIXME: this is a bit instantiation-specific.
     SemaRef.MarkMemberReferenced(E);
 
     return E;
   }
 
   TemplateArgumentListInfo TransArgs;
   if (E->hasExplicitTemplateArgs()) {
     TransArgs.setLAngleLoc(E->getLAngleLoc());
     TransArgs.setRAngleLoc(E->getRAngleLoc());
     if (getDerived().TransformTemplateArguments(E->getTemplateArgs(),
                                                 E->getNumTemplateArgs(),
                                                 TransArgs))
       return ExprError();

clang/test/OpenMP/default_firstprivate_ast_print.cpp

#pragma omp parallel default(firstprivate)
     {
       targetDev++;
     }
     // PRINT: #pragma omp parallel default(firstprivate)
     // PRINT-NEXT: {
     // PRINT-NEXT:  this->targetDev++;
     // CHECK-NEXT: }
     // DUMP: -OMPParallelDirective
-    // DUMP->NEXT: -OMPDefaultClause
+    // DUMP-NEXT: -OMPDefaultClause
+    // DUMP-NOT:   -OMPFirstprivateClause
   }
   // PRINT: template<> void apply<32U>()
   // PRINT: #pragma omp parallel default(firstprivate)
   // PRINT-NEXT: {
   // PRINT-NEXT:  this->targetDev++;
   // CHECK-NEXT: }
   // DUMP: -OMPParallelDirective
   // DUMP-NEXT: -OMPDefaultClause
+  // DUMP-NEXT:  -OMPFirstprivateClause
+  // DUMP-NEXT:    -DeclRefExpr {{.*}} 'targetDev'
 };
 
 void use_template() {
   SomeKernel aKern;
   aKern.apply<32>();
 }
 
 void foo() {
#pragma omp parallel default(firstprivate)
   // PRINT: #pragma omp parallel default(firstprivate)
   // DUMP: -OMPParallelDirective
   // DUMP-NEXT: -OMPDefaultClause
   // DUMP-NEXT: -OMPFirstprivateClause {{.*}} <implicit>
   // DUMP-NEXT: -DeclRefExpr {{.*}} 'a'
   // DUMP-NEXT: -DeclRefExpr {{.*}} 'yy'
   // DUMP-NEXT: -DeclRefExpr {{.*}} 'y'
 }
+void zoo(int);
+struct A {
+  int z;
+  int f;
+  A();
+  ~A();
+  void foo() {
+#pragma omp parallel firstprivate(z) default(firstprivate)
+    {
+      z++;
+      f++;
+      zoo(z + f);
+      f++;
+    }
+  }
+  // PRINT:  #pragma omp parallel firstprivate(this->z) default(firstprivate)
+  // DUMP:   -OMPParallelDirective
+  // DUMP-NEXT: -OMPFirstprivateClause
+  // DUMP-NEXT: -DeclRefExpr {{.*}} 'z'
+  // DUMP-NEXT: -OMPDefaultClause
+  // DUMP-NEXT: -OMPFirstprivateClause {{.*}} <implicit>
+  // DUMP-NEXT: -DeclRefExpr {{.*}} 'f'
+  // DUMP:      -CXXThisExpr {{.*}} 'A *' implicit this
+  // DUMP-NEXT: -DeclRefExpr {{.*}} 'z'
+  // DUMP-NEXT: -DeclRefExpr {{.*}} 'f'
+  void bar() {
+#pragma omp parallel firstprivate(z) default(firstprivate)
+    {
+#pragma omp parallel private(z) default(firstprivate)
+      {
+        z++;
+        f++;
+        zoo(z + f);
+        f++;
+      }
+    }
+  }
+  // PRINT:  #pragma omp parallel firstprivate(this->z) default(firstprivate)
+  // PRINT:    #pragma omp parallel private(this->z) default(firstprivate)
+  // DUMP:     -OMPParallelDirective
+  // DUMP-NEXT: -OMPFirstprivateClause
+  // DUMP-NEXT:  -DeclRefExpr {{.*}} 'z'
+  // DUMP-NEXT:  -OMPDefaultClause
+  // DUMP:        -OMPParallelDirective
+  // DUMP-NEXT:    -OMPPrivateClaus
+  // DUMP-NEXT:     -DeclRefExpr {{.*}} 'z'
+  // DUMP-NEXT:     -OMPDefaultClause
+  // DUMP-NEXT:     -OMPFirstprivateClause {{.*}} <implicit>
+  // DUMP-NEXT:      -DeclRefExpr {{.*}} 'f'
+  // DUMP:           -CXXThisExpr {{.*}} 'A *' implicit this
+  // DUMP-NEXT:      -DeclRefExpr {{.*}} 'f'
+  // DUMP:         -MemberExpr {{.*}}
+  // DUMP-NEXT:      -CXXThisExpr
+  // DUMP:       -CXXThisExpr {{.*}} 'A *' implicit this
+  // DUMP-NEXT:  -DeclRefExpr {{.*}} 'z'
+};
 #endif // HEADER

clang/test/OpenMP/default_private_ast_print.cpp

#pragma omp parallel default(private)
   }
   // PRINT: #pragma omp parallel default(private)
   // DUMP: -OMPParallelDirective
   // DUMP-NEXT: -OMPDefaultClause
   // DUMP-NEXT: -OMPPrivateClause {{.*}} <implicit>
   // DUMP-NEXT:  -DeclRefExpr {{.*}} 'a'
   // DUMP-NEXT:  -DeclRefExpr {{.*}} 'yy'
 }
+
+void zoo(int);
+struct A {
+  int z;
+  int f;
+  A();
+  ~A();
+  void foo() {
+#pragma omp parallel private(z) default(private)
+    {
+      z++;
+      f++;
+      zoo(z + f);
+      f++;
+    }
+  }
+  // PRINT:    #pragma omp parallel private(this->z) default(private)
+  // DUMP:     -OMPParallelDirective
+  // DUMP-NEXT:  -OMPPrivateClause
+  // DUMP-NEXT:    -DeclRefExpr {{.*}} 'z'
+  // DUMP-NEXT:  -OMPDefaultClause
+  // DUMP-NEXT:  -OMPPrivateClause
+  // DUMP-NEXT:    -DeclRefExpr {{.*}} 'f'
+  // DUMP:         -CXXThisExpr {{.*}} 'A *' implicit this
+  void bar() {
+#pragma omp parallel private(z) default(private)
+    {
+#pragma omp parallel private(z) default(private)
+      {
+        z++;
+        f++;
+        zoo(z + f);
+        f++;
+      }
+    }
+  }
+  // PRINT:    #pragma omp parallel private(this->z) default(private)
+  // PRINT:          #pragma omp parallel private(this->z) default(private)
+  // DUMP:     -OMPParallelDirective
+  // DUMP-NEXT:  -OMPPrivateClause
+  // DUMP-NEXT:    -DeclRefExpr {{.*}} 'z'
+  // DUMP-NEXT:  -OMPDefaultClause
+  // DUMP:           -OMPParallelDirective
+  // DUMP-NEXT:        -OMPPrivateClause
+  // DUMP-NEXT:           -DeclRefExpr {{.*}} 'z'
+  // DUMP-NEXT:        -OMPDefaultClause
+  // DUMP-NEXT:        -OMPPrivateClause {{.*}} <implicit>
+  // DUMP-NEXT:           -DeclRefExpr {{.*}} 'f'
+  // DUMP:             -CXXThisExpr
+  // DUMP:         -MemberExpr
+  // DUMP-NEXT:       -CXXThisExpr
+  // DUMP:         -CXXThisExpr
+};
 #endif // HEADER