Skip to content

Commit 4bbdb01

Browse files
erichkeaneerichkeane
committed
[OpenACC] Implement 'init' and 'shutdown' constructs
These two constructs are very simple and similar, and only support 3 different clauses, two of which are already implemented. This patch adds AST nodes for both constructs, and leaves the device_num clause unimplemented, but enables the other two.
1 parent 10d054e commit 4bbdb01

40 files changed

+810
-118
lines changed

‎clang/include/clang-c/Index.h

Lines changed: 9 additions & 1 deletion
Original file line numberDiff line numberDiff line change
@@ -2190,7 +2190,15 @@ enum CXCursorKind {
21902190
*/
21912191
CXCursor_OpenACCWaitConstruct = 327,
21922192

2193-
CXCursor_LastStmt = CXCursor_OpenACCWaitConstruct,
2193+
/** OpenACC init Construct.
2194+
*/
2195+
CXCursor_OpenACCInitConstruct = 328,
2196+
2197+
/** OpenACC shutdown Construct.
2198+
*/
2199+
CXCursor_OpenACCShutdownConstruct = 329,
2200+
2201+
CXCursor_LastStmt = CXCursor_OpenACCShutdownConstruct,
21942202

21952203
/**
21962204
* Cursor that represents the translation unit itself.

‎clang/include/clang/AST/RecursiveASTVisitor.h

Lines changed: 4 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -4076,6 +4076,10 @@ DEF_TRAVERSE_STMT(OpenACCWaitConstruct, {
40764076
TRY_TO(TraverseStmt(E));
40774077
TRY_TO(VisitOpenACCClauseList(S->clauses()));
40784078
})
4079+
DEF_TRAVERSE_STMT(OpenACCInitConstruct,
4080+
{ TRY_TO(VisitOpenACCClauseList(S->clauses())); })
4081+
DEF_TRAVERSE_STMT(OpenACCShutdownConstruct,
4082+
{ TRY_TO(VisitOpenACCClauseList(S->clauses())); })
40794083

40804084
// Traverse HLSL: Out argument expression
40814085
DEF_TRAVERSE_STMT(HLSLOutArgExpr, {})

‎clang/include/clang/AST/StmtOpenACC.h

Lines changed: 80 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -592,5 +592,85 @@ class OpenACCWaitConstruct final
592592
return const_child_range(Begin, Begin + NumExprs);
593593
}
594594
};
595+
596+
// This class represents an 'init' construct, which has just a clause list.
597+
class OpenACCInitConstruct final
598+
: public OpenACCConstructStmt,
599+
private llvm::TrailingObjects<OpenACCInitConstruct,
600+
const OpenACCClause *> {
601+
friend TrailingObjects;
602+
OpenACCInitConstruct(unsigned NumClauses)
603+
: OpenACCConstructStmt(OpenACCInitConstructClass,
604+
OpenACCDirectiveKind::Init, SourceLocation{},
605+
SourceLocation{}, SourceLocation{}) {
606+
std::uninitialized_value_construct(
607+
getTrailingObjects<const OpenACCClause *>(),
608+
getTrailingObjects<const OpenACCClause *>() + NumClauses);
609+
setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
610+
NumClauses));
611+
}
612+
OpenACCInitConstruct(SourceLocation Start, SourceLocation DirectiveLoc,
613+
SourceLocation End,
614+
ArrayRef<const OpenACCClause *> Clauses)
615+
: OpenACCConstructStmt(OpenACCInitConstructClass,
616+
OpenACCDirectiveKind::Init, Start, DirectiveLoc,
617+
End) {
618+
std::uninitialized_copy(Clauses.begin(), Clauses.end(),
619+
getTrailingObjects<const OpenACCClause *>());
620+
setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
621+
Clauses.size()));
622+
}
623+
624+
public:
625+
static bool classof(const Stmt *T) {
626+
return T->getStmtClass() == OpenACCInitConstructClass;
627+
}
628+
static OpenACCInitConstruct *CreateEmpty(const ASTContext &C,
629+
unsigned NumClauses);
630+
static OpenACCInitConstruct *Create(const ASTContext &C, SourceLocation Start,
631+
SourceLocation DirectiveLoc,
632+
SourceLocation End,
633+
ArrayRef<const OpenACCClause *> Clauses);
634+
};
635+
636+
// This class represents a 'shutdown' construct, which has just a clause list.
637+
class OpenACCShutdownConstruct final
638+
: public OpenACCConstructStmt,
639+
private llvm::TrailingObjects<OpenACCShutdownConstruct,
640+
const OpenACCClause *> {
641+
friend TrailingObjects;
642+
OpenACCShutdownConstruct(unsigned NumClauses)
643+
: OpenACCConstructStmt(OpenACCShutdownConstructClass,
644+
OpenACCDirectiveKind::Shutdown, SourceLocation{},
645+
SourceLocation{}, SourceLocation{}) {
646+
std::uninitialized_value_construct(
647+
getTrailingObjects<const OpenACCClause *>(),
648+
getTrailingObjects<const OpenACCClause *>() + NumClauses);
649+
setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
650+
NumClauses));
651+
}
652+
OpenACCShutdownConstruct(SourceLocation Start, SourceLocation DirectiveLoc,
653+
SourceLocation End,
654+
ArrayRef<const OpenACCClause *> Clauses)
655+
: OpenACCConstructStmt(OpenACCShutdownConstructClass,
656+
OpenACCDirectiveKind::Shutdown, Start,
657+
DirectiveLoc, End) {
658+
std::uninitialized_copy(Clauses.begin(), Clauses.end(),
659+
getTrailingObjects<const OpenACCClause *>());
660+
setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
661+
Clauses.size()));
662+
}
663+
664+
public:
665+
static bool classof(const Stmt *T) {
666+
return T->getStmtClass() == OpenACCShutdownConstructClass;
667+
}
668+
static OpenACCShutdownConstruct *CreateEmpty(const ASTContext &C,
669+
unsigned NumClauses);
670+
static OpenACCShutdownConstruct *
671+
Create(const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
672+
SourceLocation End, ArrayRef<const OpenACCClause *> Clauses);
673+
};
674+
595675
} // namespace clang
596676
#endif // LLVM_CLANG_AST_STMTOPENACC_H

‎clang/include/clang/AST/TextNodeDumper.h

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -416,6 +416,8 @@ class TextNodeDumper
416416
void VisitOpenACCExitDataConstruct(const OpenACCExitDataConstruct *S);
417417
void VisitOpenACCHostDataConstruct(const OpenACCHostDataConstruct *S);
418418
void VisitOpenACCWaitConstruct(const OpenACCWaitConstruct *S);
419+
void VisitOpenACCInitConstruct(const OpenACCInitConstruct *S);
420+
void VisitOpenACCShutdownConstruct(const OpenACCShutdownConstruct *S);
419421
void VisitOpenACCAsteriskSizeExpr(const OpenACCAsteriskSizeExpr *S);
420422
void VisitEmbedExpr(const EmbedExpr *S);
421423
void VisitAtomicExpr(const AtomicExpr *AE);

‎clang/include/clang/Basic/DiagnosticSemaKinds.td

Lines changed: 2 additions & 2 deletions
Original file line numberDiff line numberDiff line change
@@ -12669,8 +12669,8 @@ def err_acc_int_expr_requires_integer
1266912669
def err_acc_int_expr_incomplete_class_type
1267012670
: Error<"OpenACC integer expression has incomplete class type %0">;
1267112671
def err_acc_int_expr_explicit_conversion
12672-
: Error<"OpenACC integer expression type %0 requires explicit conversion "
12673-
"to %1">;
12672+
: Error<"OpenACC integer expression requires explicit conversion "
12673+
"from %0 to %1">;
1267412674
def note_acc_int_expr_conversion
1267512675
: Note<"conversion to %select{integral|enumeration}0 type %1">;
1267612676
def err_acc_int_expr_multiple_conversions

‎clang/include/clang/Basic/StmtNodes.td

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -313,6 +313,8 @@ def OpenACCEnterDataConstruct : StmtNode<OpenACCConstructStmt>;
313313
def OpenACCExitDataConstruct : StmtNode<OpenACCConstructStmt>;
314314
def OpenACCHostDataConstruct : StmtNode<OpenACCAssociatedStmtConstruct>;
315315
def OpenACCWaitConstruct : StmtNode<OpenACCConstructStmt>;
316+
def OpenACCInitConstruct : StmtNode<OpenACCConstructStmt>;
317+
def OpenACCShutdownConstruct : StmtNode<OpenACCConstructStmt>;
316318

317319
// OpenACC Additional Expressions.
318320
def OpenACCAsteriskSizeExpr : StmtNode<Expr>;

‎clang/include/clang/Serialization/ASTBitCodes.h

Lines changed: 2 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -2022,6 +2022,8 @@ enum StmtCode {
20222022
STMT_OPENACC_EXIT_DATA_CONSTRUCT,
20232023
STMT_OPENACC_HOST_DATA_CONSTRUCT,
20242024
STMT_OPENACC_WAIT_CONSTRUCT,
2025+
STMT_OPENACC_INIT_CONSTRUCT,
2026+
STMT_OPENACC_SHUTDOWN_CONSTRUCT,
20252027

20262028
// HLSL Constructs
20272029
EXPR_HLSL_OUT_ARG,

‎clang/lib/AST/StmtOpenACC.cpp

Lines changed: 40 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -225,3 +225,43 @@ OpenACCWaitConstruct *OpenACCWaitConstruct::Create(
225225
QueuesLoc, QueueIdExprs, RParenLoc, End, Clauses);
226226
return Inst;
227227
}
228+
OpenACCInitConstruct *OpenACCInitConstruct::CreateEmpty(const ASTContext &C,
229+
unsigned NumClauses) {
230+
void *Mem =
231+
C.Allocate(OpenACCInitConstruct::totalSizeToAlloc<const OpenACCClause *>(
232+
NumClauses));
233+
auto *Inst = new (Mem) OpenACCInitConstruct(NumClauses);
234+
return Inst;
235+
}
236+
237+
OpenACCInitConstruct *
238+
OpenACCInitConstruct::Create(const ASTContext &C, SourceLocation Start,
239+
SourceLocation DirectiveLoc, SourceLocation End,
240+
ArrayRef<const OpenACCClause *> Clauses) {
241+
void *Mem =
242+
C.Allocate(OpenACCInitConstruct::totalSizeToAlloc<const OpenACCClause *>(
243+
Clauses.size()));
244+
auto *Inst =
245+
new (Mem) OpenACCInitConstruct(Start, DirectiveLoc, End, Clauses);
246+
return Inst;
247+
}
248+
OpenACCShutdownConstruct *
249+
OpenACCShutdownConstruct::CreateEmpty(const ASTContext &C,
250+
unsigned NumClauses) {
251+
void *Mem = C.Allocate(
252+
OpenACCShutdownConstruct::totalSizeToAlloc<const OpenACCClause *>(
253+
NumClauses));
254+
auto *Inst = new (Mem) OpenACCShutdownConstruct(NumClauses);
255+
return Inst;
256+
}
257+
258+
OpenACCShutdownConstruct *OpenACCShutdownConstruct::Create(
259+
const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
260+
SourceLocation End, ArrayRef<const OpenACCClause *> Clauses) {
261+
void *Mem = C.Allocate(
262+
OpenACCShutdownConstruct::totalSizeToAlloc<const OpenACCClause *>(
263+
Clauses.size()));
264+
auto *Inst =
265+
new (Mem) OpenACCShutdownConstruct(Start, DirectiveLoc, End, Clauses);
266+
return Inst;
267+
}

‎clang/lib/AST/StmtPrinter.cpp

Lines changed: 25 additions & 62 deletions
Original file line numberDiff line numberDiff line change
@@ -127,6 +127,8 @@ namespace {
127127
void PrintOMPExecutableDirective(OMPExecutableDirective *S,
128128
bool ForceNoStmt = false);
129129
void PrintFPPragmas(CompoundStmt *S);
130+
void PrintOpenACCClauseList(OpenACCConstructStmt *S);
131+
void PrintOpenACCConstruct(OpenACCConstructStmt *S);
130132

131133
void PrintExpr(Expr *E) {
132134
if (E)
@@ -1155,87 +1157,52 @@ void StmtPrinter::VisitOMPTargetParallelGenericLoopDirective(
11551157
//===----------------------------------------------------------------------===//
11561158
// OpenACC construct printing methods
11571159
//===----------------------------------------------------------------------===//
1158-
void StmtPrinter::VisitOpenACCComputeConstruct(OpenACCComputeConstruct *S) {
1159-
Indent() << "#pragma acc " << S->getDirectiveKind();
1160-
1160+
void StmtPrinter::PrintOpenACCClauseList(OpenACCConstructStmt *S) {
11611161
if (!S->clauses().empty()) {
11621162
OS << ' ';
11631163
OpenACCClausePrinter Printer(OS, Policy);
11641164
Printer.VisitClauseList(S->clauses());
11651165
}
1166+
}
1167+
void StmtPrinter::PrintOpenACCConstruct(OpenACCConstructStmt *S) {
1168+
Indent() << "#pragma acc " << S->getDirectiveKind();
1169+
PrintOpenACCClauseList(S);
11661170
OS << '\n';
1167-
1171+
}
1172+
void StmtPrinter::VisitOpenACCComputeConstruct(OpenACCComputeConstruct *S) {
1173+
PrintOpenACCConstruct(S);
11681174
PrintStmt(S->getStructuredBlock());
11691175
}
11701176

11711177
void StmtPrinter::VisitOpenACCLoopConstruct(OpenACCLoopConstruct *S) {
1172-
Indent() << "#pragma acc loop";
1173-
1174-
if (!S->clauses().empty()) {
1175-
OS << ' ';
1176-
OpenACCClausePrinter Printer(OS, Policy);
1177-
Printer.VisitClauseList(S->clauses());
1178-
}
1179-
OS << '\n';
1180-
1178+
PrintOpenACCConstruct(S);
11811179
PrintStmt(S->getLoop());
11821180
}
11831181

11841182
void StmtPrinter::VisitOpenACCCombinedConstruct(OpenACCCombinedConstruct *S) {
1185-
Indent() << "#pragma acc " << S->getDirectiveKind();
1186-
if (!S->clauses().empty()) {
1187-
OS << ' ';
1188-
OpenACCClausePrinter Printer(OS, Policy);
1189-
Printer.VisitClauseList(S->clauses());
1190-
}
1191-
OS << '\n';
1192-
1183+
PrintOpenACCConstruct(S);
11931184
PrintStmt(S->getLoop());
11941185
}
11951186

11961187
void StmtPrinter::VisitOpenACCDataConstruct(OpenACCDataConstruct *S) {
1197-
Indent() << "#pragma acc data";
1198-
1199-
if (!S->clauses().empty()) {
1200-
OS << ' ';
1201-
OpenACCClausePrinter Printer(OS, Policy);
1202-
Printer.VisitClauseList(S->clauses());
1203-
}
1204-
OS << '\n';
1205-
1188+
PrintOpenACCConstruct(S);
1189+
PrintStmt(S->getStructuredBlock());
1190+
}
1191+
void StmtPrinter::VisitOpenACCHostDataConstruct(OpenACCHostDataConstruct *S) {
1192+
PrintOpenACCConstruct(S);
12061193
PrintStmt(S->getStructuredBlock());
12071194
}
12081195
void StmtPrinter::VisitOpenACCEnterDataConstruct(OpenACCEnterDataConstruct *S) {
1209-
Indent() << "#pragma acc enter data";
1210-
1211-
if (!S->clauses().empty()) {
1212-
OS << ' ';
1213-
OpenACCClausePrinter Printer(OS, Policy);
1214-
Printer.VisitClauseList(S->clauses());
1215-
}
1216-
OS << '\n';
1196+
PrintOpenACCConstruct(S);
12171197
}
12181198
void StmtPrinter::VisitOpenACCExitDataConstruct(OpenACCExitDataConstruct *S) {
1219-
Indent() << "#pragma acc exit data";
1220-
1221-
if (!S->clauses().empty()) {
1222-
OS << ' ';
1223-
OpenACCClausePrinter Printer(OS, Policy);
1224-
Printer.VisitClauseList(S->clauses());
1225-
}
1226-
OS << '\n';
1199+
PrintOpenACCConstruct(S);
12271200
}
1228-
void StmtPrinter::VisitOpenACCHostDataConstruct(OpenACCHostDataConstruct *S) {
1229-
Indent() << "#pragma acc host_data";
1230-
1231-
if (!S->clauses().empty()) {
1232-
OS << ' ';
1233-
OpenACCClausePrinter Printer(OS, Policy);
1234-
Printer.VisitClauseList(S->clauses());
1235-
}
1236-
OS << '\n';
1237-
1238-
PrintStmt(S->getStructuredBlock());
1201+
void StmtPrinter::VisitOpenACCInitConstruct(OpenACCInitConstruct *S) {
1202+
PrintOpenACCConstruct(S);
1203+
}
1204+
void StmtPrinter::VisitOpenACCShutdownConstruct(OpenACCShutdownConstruct *S) {
1205+
PrintOpenACCConstruct(S);
12391206
}
12401207

12411208
void StmtPrinter::VisitOpenACCWaitConstruct(OpenACCWaitConstruct *S) {
@@ -1258,11 +1225,7 @@ void StmtPrinter::VisitOpenACCWaitConstruct(OpenACCWaitConstruct *S) {
12581225
OS << ")";
12591226
}
12601227

1261-
if (!S->clauses().empty()) {
1262-
OS << ' ';
1263-
OpenACCClausePrinter Printer(OS, Policy);
1264-
Printer.VisitClauseList(S->clauses());
1265-
}
1228+
PrintOpenACCClauseList(S);
12661229
OS << '\n';
12671230
}
12681231

‎clang/lib/AST/StmtProfile.cpp

Lines changed: 13 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -2751,6 +2751,19 @@ void StmtProfiler::VisitOpenACCWaitConstruct(const OpenACCWaitConstruct *S) {
27512751
P.VisitOpenACCClauseList(S->clauses());
27522752
}
27532753

2754+
void StmtProfiler::VisitOpenACCInitConstruct(const OpenACCInitConstruct *S) {
2755+
VisitStmt(S);
2756+
OpenACCClauseProfiler P{*this};
2757+
P.VisitOpenACCClauseList(S->clauses());
2758+
}
2759+
2760+
void StmtProfiler::VisitOpenACCShutdownConstruct(
2761+
const OpenACCShutdownConstruct *S) {
2762+
VisitStmt(S);
2763+
OpenACCClauseProfiler P{*this};
2764+
P.VisitOpenACCClauseList(S->clauses());
2765+
}
2766+
27542767
void StmtProfiler::VisitHLSLOutArgExpr(const HLSLOutArgExpr *S) {
27552768
VisitStmt(S);
27562769
}

‎clang/lib/AST/TextNodeDumper.cpp

Lines changed: 7 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -2963,6 +2963,13 @@ void TextNodeDumper::VisitOpenACCHostDataConstruct(
29632963
void TextNodeDumper::VisitOpenACCWaitConstruct(const OpenACCWaitConstruct *S) {
29642964
OS << " " << S->getDirectiveKind();
29652965
}
2966+
void TextNodeDumper::VisitOpenACCInitConstruct(const OpenACCInitConstruct *S) {
2967+
OS << " " << S->getDirectiveKind();
2968+
}
2969+
void TextNodeDumper::VisitOpenACCShutdownConstruct(
2970+
const OpenACCShutdownConstruct *S) {
2971+
OS << " " << S->getDirectiveKind();
2972+
}
29662973

29672974
void TextNodeDumper::VisitEmbedExpr(const EmbedExpr *S) {
29682975
AddChild("begin", [=] { OS << S->getStartingElementPos(); });

‎clang/lib/CodeGen/CGStmt.cpp

Lines changed: 6 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -473,6 +473,12 @@ void CodeGenFunction::EmitStmt(const Stmt *S, ArrayRef<const Attr *> Attrs) {
473473
case Stmt::OpenACCWaitConstructClass:
474474
EmitOpenACCWaitConstruct(cast<OpenACCWaitConstruct>(*S));
475475
break;
476+
case Stmt::OpenACCInitConstructClass:
477+
EmitOpenACCInitConstruct(cast<OpenACCInitConstruct>(*S));
478+
break;
479+
case Stmt::OpenACCShutdownConstructClass:
480+
EmitOpenACCShutdownConstruct(cast<OpenACCShutdownConstruct>(*S));
481+
break;
476482
}
477483
}
478484

‎clang/lib/CodeGen/CodeGenFunction.h

Lines changed: 10 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -4123,6 +4123,16 @@ class CodeGenFunction : public CodeGenTypeCache {
41234123
// but in the future we will implement some sort of IR.
41244124
}
41254125

4126+
void EmitOpenACCInitConstruct(const OpenACCInitConstruct &S) {
4127+
// TODO OpenACC: Implement this. It is currently implemented as a 'no-op',
4128+
// but in the future we will implement some sort of IR.
4129+
}
4130+
4131+
void EmitOpenACCShutdownConstruct(const OpenACCShutdownConstruct &S) {
4132+
// TODO OpenACC: Implement this. It is currently implemented as a 'no-op',
4133+
// but in the future we will implement some sort of IR.
4134+
}
4135+
41264136
//===--------------------------------------------------------------------===//
41274137
// LValue Expression Emission
41284138
//===--------------------------------------------------------------------===//

0 commit comments

Comments
 (0)