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 ↗

(On Diff #438877)

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

///

1141–1142

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

1153

const auto &?

1158–1159

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

1167

const auto &?

1172

/// style

1179–1180

just emplace_back(FD, Sz, VD);?

2390

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 ↗

(On Diff #438877)

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

2390

Changed.

[ABataev]

clang/lib/Sema/SemaExprMember.cpp

1873–1877 ↗

(On Diff #438877)

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 ↗

(On Diff #438877)

Okay, I moved my code into isOpenMPCapturedDecl.

[ABataev]

clang/lib/Sema/SemaOpenMP.cpp

2421

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

2433

Just DVarPrivate.CKind != OMPC_private

clang/lib/Sema/TreeTransform.h

11120–11121

Why do we need this check here?

[jyu2]

Address Alexey's comments! Thanks.

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

2421

Thanks. Change

2433

Yes!! changed

clang/lib/Sema/TreeTransform.h

11120–11121

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

11120–11121

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

[jyu2]

clang/lib/Sema/TreeTransform.h

11120–11121

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

11120–11121

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

11120–11121

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

11120–11121

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

1149

Is it possible to have an overflow here?

1154

What if Sz == -1?

1182

What about overflow here?

17803–17813

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

18082

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

1149

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));

1154

Should not be 0.

1182

Add assert for boundary check.

17803–17813

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.

2421

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

clang/lib/Sema/TreeTransform.h

11125

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

17803–17813

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

2273–2274

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

[jyu2]

clang/lib/Sema/SemaOpenMP.cpp

2273–2274

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

2273–2274

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

2273–2274

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);
+
+  /// The member expression(this->fd) needs to be rebuilt in the template
+  /// instantiation to generate private copy for OpenMP when default
+  /// clause is used. The function will return true if default
+  /// cluse is used.
+  bool isOpenMPRebuildMemberExpr(ValueDecl *D);
+
   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/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;
+    /// Data is required on creating capture fields for implicit
+    /// default first|private clause.
+    struct ImplicitDefaultFDInfoTy {
+      /// Field decl.
+      const FieldDecl *FD = nullptr;
+      /// Nesting stack level
+      size_t StackLevel = 0;
+      /// Capture variable decl.
+      VarDecl *VD = nullptr;
+      ImplicitDefaultFDInfoTy(const FieldDecl *FD, size_t StackLevel,
+                              VarDecl *VD)
+          : FD(FD), StackLevel(StackLevel), VD(VD) {}
+    };
+    /// List of captured 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;
   };
 
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 StackLevel = getStackSize();
+    for (; I != EndI; ++I) {
+      if (I->DefaultAttr == DSA_firstprivate || I->DefaultAttr == DSA_private)
+        break;
+      StackLevel--;
+    }
+    assert((StackLevel > 0 && I != EndI) || (StackLevel == 0 && I == EndI));
+    if (I == EndI)
+      return nullptr;
+    for (const auto &IFD : I->ImplicitDefaultFirstprivateFDs)
+      if (IFD.FD == FD && IFD.StackLevel == StackLevel)
+        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 StackLevel = getStackSize();
+    for (; I != EndI; ++I) {
+      if (I->DefaultAttr == DSA_private || I->DefaultAttr == DSA_firstprivate) {
+        I->ImplicitDefaultFirstprivateFDs.emplace_back(FD, StackLevel, VD);
+        break;
+      }
+      StackLevel--;
+    }
+    assert((StackLevel > 0 && I != EndI) || (StackLevel == 0 && I == EndI));
+  }
 };
 
 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);
 }
 
+bool Sema::isOpenMPRebuildMemberExpr(ValueDecl *D) {
+  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 true;
+  return false;
+}
+
+static OMPCapturedExprDecl *buildCaptureDecl(Sema &S, IdentifierInfo *Id,
+                                             Expr *CaptureExpr, bool WithInit,
+                                             DeclContext *CurContext,
+                                             bool AsExpression);
+
 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;
+    auto *FD = dyn_cast<FieldDecl>(D);
+    if (DVarPrivate.CKind != OMPC_unknown && !VD && FD &&
+        !DVarPrivate.PrivateCopy) {
+      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)
+        return VD;
+      if (getCurrentThisType().isNull())
+        return nullptr;
+      Expr *ThisExpr = BuildCXXThisExpr(SourceLocation(), getCurrentThisType(),
+                                        /*IsImplicit=*/true);
+      const CXXScopeSpec CS = CXXScopeSpec();
+      Expr *ME = BuildMemberExpr(ThisExpr, /*IsArrow=*/true, SourceLocation(),
+                                 NestedNameSpecifierLoc(), SourceLocation(), FD,
+                                 DeclAccessPair::make(FD, FD->getAccess()),
+                                 /*HadMultipleCandidates=*/false,
+                                 DeclarationNameInfo(), FD->getType(),
+                                 VK_LValue, OK_Ordinary);
+      OMPCapturedExprDecl *CD = buildCaptureDecl(
+          *this, FD->getIdentifier(), ME, DVarPrivate.CKind != OMPC_private,
+          CurContext->getParent(), /*AsExpression=*/false);
+      DeclRefExpr *VDPrivateRefExpr = buildDeclRefExpr(
+          *this, CD, CD->getType().getNonReferenceType(), SourceLocation());
+      VD = cast<VarDecl>(VDPrivateRefExpr->getDecl());
+      DSAStack->addImplicitDefaultFirstprivateFD(FD, VD);
+      return VD;
+    }
     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 nullptr;
 }
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.
       if (DVar.RefExpr || !ImplicitDeclarations.insert(VD).second)
         return;
 
       // Skip internally declared static variables.
       llvm::Optional<OMPDeclareTargetDeclAttr::MapTypeTy> Res =
           OMPDeclareTargetDeclAttr::isDeclareTargetDeclaration(VD);
       if (VD->hasGlobalStorage() && CS && !CS->capturesVariable(VD) &&
           (Stack->hasRequiresDeclWithClause<OMPUnifiedSharedMemoryClause>() ||
            !Res || *Res != OMPDeclareTargetDeclAttr::MT_Link) &&
+          !Stack->isImplicitDefaultFirstprivateFD(VD) &&
           !Stack->isImplicitTaskFirstprivate(VD))
         return;
 
       SourceLocation ELoc = E->getExprLoc();
       OpenMPDirectiveKind DKind = Stack->getCurrentDirective();
       // The default(none) clause requires that each variable that is referenced
       // in the construct, and does not have a predetermined data-sharing
       // attribute, must have its data-sharing attribute explicitly determined

 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))
+        if (VD || !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()
                        : Ref);
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()) {
     if (!FoundDecl)
       return ExprError();
   }
 
   if (!getDerived().AlwaysRebuild() &&
       Base.get() == E->getBase() &&
       QualifierLoc == E->getQualifierLoc() &&
       Member == E->getMemberDecl() &&
       FoundDecl == E->getFoundDecl() &&
       !E->hasExplicitTemplateArgs()) {
 
+    // Skip for member expression of (this->f), rebuilt thisi->f is needed
+    // for Openmp where the field need to be privatizized in the case.
+    if (!(isa<CXXThisExpr>(E->getBase()) &&
+          getSema().isOpenMPRebuildMemberExpr(cast<ValueDecl>(Member)))) {
       // 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))

clang/test/OpenMP/default_firstprivate_ast_print.cpp

struct SomeKernel {
   float devPtr;
   SomeKernel();
   ~SomeKernel();
 
   template <unsigned int nRHS>
   void apply() {
 #pragma omp parallel default(firstprivate)
     {
-      targetDev++;
+      [=]() -> int {
+        return targetDev++;
+      }();
     }
     // PRINT: #pragma omp parallel default(firstprivate)
     // PRINT-NEXT: {
-    // PRINT-NEXT:  this->targetDev++;
-    // CHECK-NEXT: }
+    // PRINT-NEXT:  [=]() -> int {
+    // PRINT-NEXT:     return this->targetDev++;
+    // PRINT-NEXT:  }();
+    // PRINT-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++;
+  // PRINT-NEXT:  [=]() -> int {
+  // PRINT-NEXT:     return this->targetDev++;
+  // PRINT-NEXT:  }();
   // 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