Skip to content

Commit 64ff52a

Browse files
kevinsalakparzysz
andauthored
[OpenMP][Clang] Add parser/semantic support for dyn_groupprivate clause (llvm#152651)
This PR adds support for the `dyn_groupprivate` clause, which will be part of OpenMP 6.1. This feature allows users to request dynamic shared memory on target regions. --------- Co-authored-by: Krzysztof Parzyszek <[email protected]>
1 parent 3511388 commit 64ff52a

File tree

20 files changed

+718
-7
lines changed

20 files changed

+718
-7
lines changed

clang/include/clang/AST/OpenMPClause.h

Lines changed: 146 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -10068,6 +10068,152 @@ class OMPXDynCGroupMemClause
1006810068
Expr *getSize() const { return getStmtAs<Expr>(); }
1006910069
};
1007010070

10071+
/// This represents 'dyn_groupprivate' clause in '#pragma omp target ...'
10072+
/// and '#pragma omp teams ...' directives.
10073+
///
10074+
/// \code
10075+
/// #pragma omp target [...] dyn_groupprivate(a,b: N)
10076+
/// \endcode
10077+
class OMPDynGroupprivateClause : public OMPClause, public OMPClauseWithPreInit {
10078+
friend class OMPClauseReader;
10079+
10080+
/// Location of '('.
10081+
SourceLocation LParenLoc;
10082+
10083+
/// Modifiers for 'dyn_groupprivate' clause.
10084+
enum { SIMPLE, FALLBACK, NUM_MODIFIERS };
10085+
unsigned Modifiers[NUM_MODIFIERS];
10086+
10087+
/// Locations of modifiers.
10088+
SourceLocation ModifiersLoc[NUM_MODIFIERS];
10089+
10090+
/// The size of the dyn_groupprivate.
10091+
Expr *Size = nullptr;
10092+
10093+
/// Set the first dyn_groupprivate modifier.
10094+
///
10095+
/// \param M The modifier.
10096+
void setDynGroupprivateModifier(OpenMPDynGroupprivateClauseModifier M) {
10097+
Modifiers[SIMPLE] = M;
10098+
}
10099+
10100+
/// Set the second dyn_groupprivate modifier.
10101+
///
10102+
/// \param M The modifier.
10103+
void setDynGroupprivateFallbackModifier(
10104+
OpenMPDynGroupprivateClauseFallbackModifier M) {
10105+
Modifiers[FALLBACK] = M;
10106+
}
10107+
10108+
/// Set location of the first dyn_groupprivate modifier.
10109+
void setDynGroupprivateModifierLoc(SourceLocation Loc) {
10110+
ModifiersLoc[SIMPLE] = Loc;
10111+
}
10112+
10113+
/// Set location of the second dyn_groupprivate modifier.
10114+
void setDynGroupprivateFallbackModifierLoc(SourceLocation Loc) {
10115+
ModifiersLoc[FALLBACK] = Loc;
10116+
}
10117+
10118+
/// Sets the location of '('.
10119+
///
10120+
/// \param Loc Location of '('.
10121+
void setLParenLoc(SourceLocation Loc) { LParenLoc = Loc; }
10122+
10123+
/// Set size.
10124+
///
10125+
/// \param E Size.
10126+
void setSize(Expr *E) { Size = E; }
10127+
10128+
public:
10129+
/// Build 'dyn_groupprivate' clause with a size expression \a Size.
10130+
///
10131+
/// \param StartLoc Starting location of the clause.
10132+
/// \param LParenLoc Location of '('.
10133+
/// \param EndLoc Ending location of the clause.
10134+
/// \param Size Size.
10135+
/// \param M1 The first modifier applied to 'dyn_groupprivate' clause.
10136+
/// \param M1Loc Location of the first modifier.
10137+
/// \param M2 The second modifier applied to 'dyn_groupprivate' clause.
10138+
/// \param M2Loc Location of the second modifier.
10139+
OMPDynGroupprivateClause(SourceLocation StartLoc, SourceLocation LParenLoc,
10140+
SourceLocation EndLoc, Expr *Size, Stmt *HelperSize,
10141+
OpenMPDirectiveKind CaptureRegion,
10142+
OpenMPDynGroupprivateClauseModifier M1,
10143+
SourceLocation M1Loc,
10144+
OpenMPDynGroupprivateClauseFallbackModifier M2,
10145+
SourceLocation M2Loc)
10146+
: OMPClause(llvm::omp::OMPC_dyn_groupprivate, StartLoc, EndLoc),
10147+
OMPClauseWithPreInit(this), LParenLoc(LParenLoc), Size(Size) {
10148+
setPreInitStmt(HelperSize, CaptureRegion);
10149+
Modifiers[SIMPLE] = M1;
10150+
Modifiers[FALLBACK] = M2;
10151+
ModifiersLoc[SIMPLE] = M1Loc;
10152+
ModifiersLoc[FALLBACK] = M2Loc;
10153+
}
10154+
10155+
/// Build an empty clause.
10156+
explicit OMPDynGroupprivateClause()
10157+
: OMPClause(llvm::omp::OMPC_dyn_groupprivate, SourceLocation(),
10158+
SourceLocation()),
10159+
OMPClauseWithPreInit(this) {
10160+
Modifiers[SIMPLE] = OMPC_DYN_GROUPPRIVATE_unknown;
10161+
Modifiers[FALLBACK] = OMPC_DYN_GROUPPRIVATE_FALLBACK_unknown;
10162+
}
10163+
10164+
/// Get the first modifier of the clause.
10165+
OpenMPDynGroupprivateClauseModifier getDynGroupprivateModifier() const {
10166+
return static_cast<OpenMPDynGroupprivateClauseModifier>(Modifiers[SIMPLE]);
10167+
}
10168+
10169+
/// Get the second modifier of the clause.
10170+
OpenMPDynGroupprivateClauseFallbackModifier
10171+
getDynGroupprivateFallbackModifier() const {
10172+
return static_cast<OpenMPDynGroupprivateClauseFallbackModifier>(
10173+
Modifiers[FALLBACK]);
10174+
}
10175+
10176+
/// Get location of '('.
10177+
SourceLocation getLParenLoc() { return LParenLoc; }
10178+
10179+
/// Get the first modifier location.
10180+
SourceLocation getDynGroupprivateModifierLoc() const {
10181+
return ModifiersLoc[SIMPLE];
10182+
}
10183+
10184+
/// Get the second modifier location.
10185+
SourceLocation getDynGroupprivateFallbackModifierLoc() const {
10186+
return ModifiersLoc[FALLBACK];
10187+
}
10188+
10189+
/// Get size.
10190+
Expr *getSize() { return Size; }
10191+
10192+
/// Get size.
10193+
const Expr *getSize() const { return Size; }
10194+
10195+
child_range children() {
10196+
return child_range(reinterpret_cast<Stmt **>(&Size),
10197+
reinterpret_cast<Stmt **>(&Size) + 1);
10198+
}
10199+
10200+
const_child_range children() const {
10201+
auto Children = const_cast<OMPDynGroupprivateClause *>(this)->children();
10202+
return const_child_range(Children.begin(), Children.end());
10203+
}
10204+
10205+
child_range used_children() {
10206+
return child_range(child_iterator(), child_iterator());
10207+
}
10208+
const_child_range used_children() const {
10209+
return const_child_range(const_child_iterator(), const_child_iterator());
10210+
}
10211+
10212+
static bool classof(const OMPClause *T) {
10213+
return T->getClauseKind() == llvm::omp::OMPC_dyn_groupprivate;
10214+
}
10215+
};
10216+
1007110217
/// This represents the 'doacross' clause for the '#pragma omp ordered'
1007210218
/// directive.
1007310219
///

clang/include/clang/AST/RecursiveASTVisitor.h

Lines changed: 8 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -4165,6 +4165,14 @@ bool RecursiveASTVisitor<Derived>::VisitOMPXDynCGroupMemClause(
41654165
return true;
41664166
}
41674167

4168+
template <typename Derived>
4169+
bool RecursiveASTVisitor<Derived>::VisitOMPDynGroupprivateClause(
4170+
OMPDynGroupprivateClause *C) {
4171+
TRY_TO(VisitOMPClauseWithPreInit(C));
4172+
TRY_TO(TraverseStmt(C->getSize()));
4173+
return true;
4174+
}
4175+
41684176
template <typename Derived>
41694177
bool RecursiveASTVisitor<Derived>::VisitOMPDoacrossClause(
41704178
OMPDoacrossClause *C) {

clang/include/clang/Basic/DiagnosticSemaKinds.td

Lines changed: 3 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -12137,6 +12137,9 @@ def err_omp_unexpected_schedule_modifier : Error<
1213712137
"modifier '%0' cannot be used along with modifier '%1'">;
1213812138
def err_omp_schedule_nonmonotonic_static : Error<
1213912139
"'nonmonotonic' modifier can only be specified with 'dynamic' or 'guided' schedule kind">;
12140+
def err_omp_incompatible_dyn_groupprivate_modifier
12141+
: Error<"modifier '%0' cannot be used along with modifier '%1' in "
12142+
"dyn_groupprivate">;
1214012143
def err_omp_simple_clause_incompatible_with_ordered : Error<
1214112144
"'%0' clause with '%1' modifier cannot be specified if an 'ordered' clause is specified">;
1214212145
def err_omp_ordered_simd : Error<

clang/include/clang/Basic/OpenMPKinds.def

Lines changed: 16 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -86,6 +86,12 @@
8686
#ifndef OPENMP_GRAINSIZE_MODIFIER
8787
#define OPENMP_GRAINSIZE_MODIFIER(Name)
8888
#endif
89+
#ifndef OPENMP_DYN_GROUPPRIVATE_MODIFIER
90+
#define OPENMP_DYN_GROUPPRIVATE_MODIFIER(Name)
91+
#endif
92+
#ifndef OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER
93+
#define OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER(Name)
94+
#endif
8995
#ifndef OPENMP_NUMTASKS_MODIFIER
9096
#define OPENMP_NUMTASKS_MODIFIER(Name)
9197
#endif
@@ -242,6 +248,14 @@ OPENMP_BIND_KIND(thread)
242248
// Modifiers for the 'grainsize' clause.
243249
OPENMP_GRAINSIZE_MODIFIER(strict)
244250

251+
// Modifiers for the 'dyn_groupprivate' clause.
252+
OPENMP_DYN_GROUPPRIVATE_MODIFIER(cgroup)
253+
254+
// Fallback modifiers for the 'dyn_groupprivate' clause.
255+
OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER(abort)
256+
OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER(null)
257+
OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER(default_mem)
258+
245259
// Modifiers for the 'num_tasks' clause.
246260
OPENMP_NUMTASKS_MODIFIER(strict)
247261

@@ -263,6 +277,8 @@ OPENMP_THREADSET_KIND(omp_team)
263277

264278
#undef OPENMP_NUMTASKS_MODIFIER
265279
#undef OPENMP_NUMTHREADS_MODIFIER
280+
#undef OPENMP_DYN_GROUPPRIVATE_MODIFIER
281+
#undef OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER
266282
#undef OPENMP_GRAINSIZE_MODIFIER
267283
#undef OPENMP_BIND_KIND
268284
#undef OPENMP_ADJUST_ARGS_KIND

clang/include/clang/Basic/OpenMPKinds.h

Lines changed: 14 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -224,6 +224,20 @@ enum OpenMPGrainsizeClauseModifier {
224224
OMPC_GRAINSIZE_unknown
225225
};
226226

227+
enum OpenMPDynGroupprivateClauseModifier {
228+
#define OPENMP_DYN_GROUPPRIVATE_MODIFIER(Name) OMPC_DYN_GROUPPRIVATE_##Name,
229+
#include "clang/Basic/OpenMPKinds.def"
230+
OMPC_DYN_GROUPPRIVATE_unknown
231+
};
232+
233+
enum OpenMPDynGroupprivateClauseFallbackModifier {
234+
OMPC_DYN_GROUPPRIVATE_FALLBACK_unknown = OMPC_DYN_GROUPPRIVATE_unknown,
235+
#define OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER(Name) \
236+
OMPC_DYN_GROUPPRIVATE_FALLBACK_##Name,
237+
#include "clang/Basic/OpenMPKinds.def"
238+
OMPC_DYN_GROUPPRIVATE_FALLBACK_last
239+
};
240+
227241
enum OpenMPNumTasksClauseModifier {
228242
#define OPENMP_NUMTASKS_MODIFIER(Name) OMPC_NUMTASKS_##Name,
229243
#include "clang/Basic/OpenMPKinds.def"

clang/include/clang/Sema/SemaOpenMP.h

Lines changed: 7 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -1411,6 +1411,13 @@ class SemaOpenMP : public SemaBase {
14111411
SourceLocation LParenLoc,
14121412
SourceLocation EndLoc);
14131413

1414+
/// Called on a well-formed 'dyn_groupprivate' clause.
1415+
OMPClause *ActOnOpenMPDynGroupprivateClause(
1416+
OpenMPDynGroupprivateClauseModifier M1,
1417+
OpenMPDynGroupprivateClauseFallbackModifier M2, Expr *Size,
1418+
SourceLocation StartLoc, SourceLocation LParenLoc, SourceLocation M1Loc,
1419+
SourceLocation M2Loc, SourceLocation EndLoc);
1420+
14141421
/// Called on well-formed 'doacross' clause.
14151422
OMPClause *
14161423
ActOnOpenMPDoacrossClause(OpenMPDoacrossClauseModifier DepType,

clang/lib/AST/OpenMPClause.cpp

Lines changed: 20 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -105,6 +105,8 @@ const OMPClauseWithPreInit *OMPClauseWithPreInit::get(const OMPClause *C) {
105105
return static_cast<const OMPFilterClause *>(C);
106106
case OMPC_ompx_dyn_cgroup_mem:
107107
return static_cast<const OMPXDynCGroupMemClause *>(C);
108+
case OMPC_dyn_groupprivate:
109+
return static_cast<const OMPDynGroupprivateClause *>(C);
108110
case OMPC_message:
109111
return static_cast<const OMPMessageClause *>(C);
110112
case OMPC_default:
@@ -2857,6 +2859,24 @@ void OMPClausePrinter::VisitOMPXDynCGroupMemClause(
28572859
OS << ")";
28582860
}
28592861

2862+
void OMPClausePrinter::VisitOMPDynGroupprivateClause(
2863+
OMPDynGroupprivateClause *Node) {
2864+
OS << "dyn_groupprivate(";
2865+
if (Node->getDynGroupprivateModifier() != OMPC_DYN_GROUPPRIVATE_unknown) {
2866+
OS << getOpenMPSimpleClauseTypeName(OMPC_dyn_groupprivate,
2867+
Node->getDynGroupprivateModifier());
2868+
if (Node->getDynGroupprivateFallbackModifier() !=
2869+
OMPC_DYN_GROUPPRIVATE_FALLBACK_unknown) {
2870+
OS << ", ";
2871+
OS << getOpenMPSimpleClauseTypeName(
2872+
OMPC_dyn_groupprivate, Node->getDynGroupprivateFallbackModifier());
2873+
}
2874+
OS << ": ";
2875+
}
2876+
Node->getSize()->printPretty(OS, nullptr, Policy, 0);
2877+
OS << ')';
2878+
}
2879+
28602880
void OMPClausePrinter::VisitOMPDoacrossClause(OMPDoacrossClause *Node) {
28612881
OS << "doacross(";
28622882
OpenMPDoacrossClauseModifier DepType = Node->getDependenceType();

clang/lib/AST/StmtProfile.cpp

Lines changed: 6 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -968,6 +968,12 @@ void OMPClauseProfiler::VisitOMPXDynCGroupMemClause(
968968
if (Expr *Size = C->getSize())
969969
Profiler->VisitStmt(Size);
970970
}
971+
void OMPClauseProfiler::VisitOMPDynGroupprivateClause(
972+
const OMPDynGroupprivateClause *C) {
973+
VisitOMPClauseWithPreInit(C);
974+
if (auto *Size = C->getSize())
975+
Profiler->VisitStmt(Size);
976+
}
971977
void OMPClauseProfiler::VisitOMPDoacrossClause(const OMPDoacrossClause *C) {
972978
VisitOMPClauseList(C);
973979
}

clang/lib/Basic/OpenMPKinds.cpp

Lines changed: 24 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -196,6 +196,16 @@ unsigned clang::getOpenMPSimpleClauseType(OpenMPClauseKind Kind, StringRef Str,
196196
return OMPC_GRAINSIZE_unknown;
197197
return Type;
198198
}
199+
case OMPC_dyn_groupprivate: {
200+
return llvm::StringSwitch<unsigned>(Str)
201+
#define OPENMP_DYN_GROUPPRIVATE_MODIFIER(Name) \
202+
.Case(#Name, OMPC_DYN_GROUPPRIVATE_##Name)
203+
#define OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER(Name) \
204+
.Case(#Name, OMPC_DYN_GROUPPRIVATE_FALLBACK_##Name) \
205+
.Case("fallback(" #Name ")", OMPC_DYN_GROUPPRIVATE_FALLBACK_##Name)
206+
#include "clang/Basic/OpenMPKinds.def"
207+
.Default(OMPC_DYN_GROUPPRIVATE_unknown);
208+
}
199209
case OMPC_num_tasks: {
200210
unsigned Type = llvm::StringSwitch<unsigned>(Str)
201211
#define OPENMP_NUMTASKS_MODIFIER(Name) .Case(#Name, OMPC_NUMTASKS_##Name)
@@ -544,6 +554,20 @@ const char *clang::getOpenMPSimpleClauseTypeName(OpenMPClauseKind Kind,
544554
#include "clang/Basic/OpenMPKinds.def"
545555
}
546556
llvm_unreachable("Invalid OpenMP 'grainsize' clause modifier");
557+
case OMPC_dyn_groupprivate:
558+
switch (Type) {
559+
case OMPC_DYN_GROUPPRIVATE_unknown:
560+
case OMPC_DYN_GROUPPRIVATE_FALLBACK_last:
561+
return "unknown";
562+
#define OPENMP_DYN_GROUPPRIVATE_MODIFIER(Name) \
563+
case OMPC_DYN_GROUPPRIVATE_##Name: \
564+
return #Name;
565+
#define OPENMP_DYN_GROUPPRIVATE_FALLBACK_MODIFIER(Name) \
566+
case OMPC_DYN_GROUPPRIVATE_FALLBACK_##Name: \
567+
return "fallback(" #Name ")";
568+
#include "clang/Basic/OpenMPKinds.def"
569+
}
570+
llvm_unreachable("Invalid OpenMP 'dyn_groupprivate' clause modifier");
547571
case OMPC_num_tasks:
548572
switch (Type) {
549573
case OMPC_NUMTASKS_unknown:

0 commit comments

Comments
 (0)