[clang] 010d011 - [OpenACC] Create AST nodes for 'data' constructs
via cfe-commits
cfe-commits at lists.llvm.org
Thu Dec 12 07:28:35 PST 2024
Author: erichkeane
Date: 2024-12-12T07:28:30-08:00
New Revision: 010d0115fc8e3834fc6f747f0841f3b1e467c4da
URL: https://github.com/llvm/llvm-project/commit/010d0115fc8e3834fc6f747f0841f3b1e467c4da
DIFF: https://github.com/llvm/llvm-project/commit/010d0115fc8e3834fc6f747f0841f3b1e467c4da.diff
LOG: [OpenACC] Create AST nodes for 'data' constructs
These constructs are all very similar and closely related, so this patch
creates the AST nodes for them, serialization, printing/etc.
Additionally the restrictions are all added as tests/todos in the tests,
as those will have to be implemented once we get those clauses implemented.
Added:
clang/test/AST/ast-print-openacc-data-construct.cpp
clang/test/SemaOpenACC/data-construct-ast.cpp
clang/test/SemaOpenACC/data-construct.cpp
Modified:
clang/include/clang-c/Index.h
clang/include/clang/AST/RecursiveASTVisitor.h
clang/include/clang/AST/StmtOpenACC.h
clang/include/clang/AST/TextNodeDumper.h
clang/include/clang/Basic/StmtNodes.td
clang/include/clang/Serialization/ASTBitCodes.h
clang/lib/AST/StmtOpenACC.cpp
clang/lib/AST/StmtPrinter.cpp
clang/lib/AST/StmtProfile.cpp
clang/lib/AST/TextNodeDumper.cpp
clang/lib/CodeGen/CGStmt.cpp
clang/lib/CodeGen/CodeGenFunction.h
clang/lib/Parse/ParseOpenACC.cpp
clang/lib/Sema/SemaExceptionSpec.cpp
clang/lib/Sema/SemaOpenACC.cpp
clang/lib/Sema/TreeTransform.h
clang/lib/Serialization/ASTReaderStmt.cpp
clang/lib/Serialization/ASTWriterStmt.cpp
clang/lib/StaticAnalyzer/Core/ExprEngine.cpp
clang/test/ParserOpenACC/parse-clauses.c
clang/test/ParserOpenACC/parse-clauses.cpp
clang/test/ParserOpenACC/parse-constructs.c
clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp
clang/test/SemaOpenACC/combined-construct-default-clause.c
clang/test/SemaOpenACC/combined-construct-if-clause.c
clang/test/SemaOpenACC/compute-construct-default-clause.c
clang/test/SemaOpenACC/compute-construct-device_type-clause.c
clang/test/SemaOpenACC/compute-construct-if-clause.c
clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
clang/tools/libclang/CIndex.cpp
clang/tools/libclang/CXCursor.cpp
Removed:
################################################################################
diff --git a/clang/include/clang-c/Index.h b/clang/include/clang-c/Index.h
index 8fc06328f0bcef..29858f00fad74e 100644
--- a/clang/include/clang-c/Index.h
+++ b/clang/include/clang-c/Index.h
@@ -2166,9 +2166,27 @@ enum CXCursorKind {
*/
CXCursor_OpenACCLoopConstruct = 321,
+ /** OpenACC Combined Constructs.
+ */
CXCursor_OpenACCCombinedConstruct = 322,
- CXCursor_LastStmt = CXCursor_OpenACCCombinedConstruct,
+ /** OpenACC data Construct.
+ */
+ CXCursor_OpenACCDataConstruct = 323,
+
+ /** OpenACC enter data Construct.
+ */
+ CXCursor_OpenACCEnterDataConstruct = 324,
+
+ /** OpenACC exit data Construct.
+ */
+ CXCursor_OpenACCExitDataConstruct = 325,
+
+ /** OpenACC host_data Construct.
+ */
+ CXCursor_OpenACCHostDataConstruct = 326,
+
+ CXCursor_LastStmt = CXCursor_OpenACCHostDataConstruct,
/**
* Cursor that represents the translation unit itself.
diff --git a/clang/include/clang/AST/RecursiveASTVisitor.h b/clang/include/clang/AST/RecursiveASTVisitor.h
index 76b598a5db2382..33363072c716f2 100644
--- a/clang/include/clang/AST/RecursiveASTVisitor.h
+++ b/clang/include/clang/AST/RecursiveASTVisitor.h
@@ -4058,6 +4058,12 @@ DEF_TRAVERSE_STMT(OpenACCLoopConstruct,
{ TRY_TO(TraverseOpenACCAssociatedStmtConstruct(S)); })
DEF_TRAVERSE_STMT(OpenACCCombinedConstruct,
{ TRY_TO(TraverseOpenACCAssociatedStmtConstruct(S)); })
+DEF_TRAVERSE_STMT(OpenACCDataConstruct,
+ { TRY_TO(TraverseOpenACCAssociatedStmtConstruct(S)); })
+DEF_TRAVERSE_STMT(OpenACCEnterDataConstruct, {})
+DEF_TRAVERSE_STMT(OpenACCExitDataConstruct, {})
+DEF_TRAVERSE_STMT(OpenACCHostDataConstruct,
+ { TRY_TO(TraverseOpenACCAssociatedStmtConstruct(S)); })
// Traverse HLSL: Out argument expression
DEF_TRAVERSE_STMT(HLSLOutArgExpr, {})
diff --git a/clang/include/clang/AST/StmtOpenACC.h b/clang/include/clang/AST/StmtOpenACC.h
index fa8793e740822f..df73980822c7be 100644
--- a/clang/include/clang/AST/StmtOpenACC.h
+++ b/clang/include/clang/AST/StmtOpenACC.h
@@ -292,5 +292,175 @@ class OpenACCCombinedConstruct final
return const_cast<OpenACCCombinedConstruct *>(this)->getLoop();
}
};
+
+// This class represents a 'data' construct, which has an associated statement
+// and clauses, but is otherwise pretty simple.
+class OpenACCDataConstruct final
+ : public OpenACCAssociatedStmtConstruct,
+ public llvm::TrailingObjects<OpenACCCombinedConstruct,
+ const OpenACCClause *> {
+ OpenACCDataConstruct(unsigned NumClauses)
+ : OpenACCAssociatedStmtConstruct(
+ OpenACCDataConstructClass, OpenACCDirectiveKind::Data,
+ SourceLocation{}, SourceLocation{}, SourceLocation{},
+ /*AssociatedStmt=*/nullptr) {
+ std::uninitialized_value_construct(
+ getTrailingObjects<const OpenACCClause *>(),
+ getTrailingObjects<const OpenACCClause *>() + NumClauses);
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ NumClauses));
+ }
+
+ OpenACCDataConstruct(SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End,
+ ArrayRef<const OpenACCClause *> Clauses,
+ Stmt *StructuredBlock)
+ : OpenACCAssociatedStmtConstruct(OpenACCDataConstructClass,
+ OpenACCDirectiveKind::Data, Start,
+ DirectiveLoc, End, StructuredBlock) {
+ std::uninitialized_copy(Clauses.begin(), Clauses.end(),
+ getTrailingObjects<const OpenACCClause *>());
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ Clauses.size()));
+ }
+ void setStructuredBlock(Stmt *S) { setAssociatedStmt(S); }
+
+public:
+ static bool classof(const Stmt *T) {
+ return T->getStmtClass() == OpenACCDataConstructClass;
+ }
+
+ static OpenACCDataConstruct *CreateEmpty(const ASTContext &C,
+ unsigned NumClauses);
+ static OpenACCDataConstruct *Create(const ASTContext &C, SourceLocation Start,
+ SourceLocation DirectiveLoc,
+ SourceLocation End,
+ ArrayRef<const OpenACCClause *> Clauses,
+ Stmt *StructuredBlock);
+ Stmt *getStructuredBlock() { return getAssociatedStmt(); }
+ const Stmt *getStructuredBlock() const {
+ return const_cast<OpenACCDataConstruct *>(this)->getStructuredBlock();
+ }
+};
+// This class represents a 'enter data' construct, which JUST has clauses.
+class OpenACCEnterDataConstruct final
+ : public OpenACCConstructStmt,
+ public llvm::TrailingObjects<OpenACCCombinedConstruct,
+ const OpenACCClause *> {
+ OpenACCEnterDataConstruct(unsigned NumClauses)
+ : OpenACCConstructStmt(OpenACCEnterDataConstructClass,
+ OpenACCDirectiveKind::EnterData, SourceLocation{},
+ SourceLocation{}, SourceLocation{}) {
+ std::uninitialized_value_construct(
+ getTrailingObjects<const OpenACCClause *>(),
+ getTrailingObjects<const OpenACCClause *>() + NumClauses);
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ NumClauses));
+ }
+ OpenACCEnterDataConstruct(SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End,
+ ArrayRef<const OpenACCClause *> Clauses)
+ : OpenACCConstructStmt(OpenACCEnterDataConstructClass,
+ OpenACCDirectiveKind::EnterData, Start,
+ DirectiveLoc, End) {
+ std::uninitialized_copy(Clauses.begin(), Clauses.end(),
+ getTrailingObjects<const OpenACCClause *>());
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ Clauses.size()));
+ }
+
+public:
+ static bool classof(const Stmt *T) {
+ return T->getStmtClass() == OpenACCEnterDataConstructClass;
+ }
+ static OpenACCEnterDataConstruct *CreateEmpty(const ASTContext &C,
+ unsigned NumClauses);
+ static OpenACCEnterDataConstruct *
+ Create(const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End, ArrayRef<const OpenACCClause *> Clauses);
+};
+// This class represents a 'exit data' construct, which JUST has clauses.
+class OpenACCExitDataConstruct final
+ : public OpenACCConstructStmt,
+ public llvm::TrailingObjects<OpenACCCombinedConstruct,
+ const OpenACCClause *> {
+ OpenACCExitDataConstruct(unsigned NumClauses)
+ : OpenACCConstructStmt(OpenACCExitDataConstructClass,
+ OpenACCDirectiveKind::ExitData, SourceLocation{},
+ SourceLocation{}, SourceLocation{}) {
+ std::uninitialized_value_construct(
+ getTrailingObjects<const OpenACCClause *>(),
+ getTrailingObjects<const OpenACCClause *>() + NumClauses);
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ NumClauses));
+ }
+ OpenACCExitDataConstruct(SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End,
+ ArrayRef<const OpenACCClause *> Clauses)
+ : OpenACCConstructStmt(OpenACCExitDataConstructClass,
+ OpenACCDirectiveKind::ExitData, Start,
+ DirectiveLoc, End) {
+ std::uninitialized_copy(Clauses.begin(), Clauses.end(),
+ getTrailingObjects<const OpenACCClause *>());
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ Clauses.size()));
+ }
+
+public:
+ static bool classof(const Stmt *T) {
+ return T->getStmtClass() == OpenACCExitDataConstructClass;
+ }
+ static OpenACCExitDataConstruct *CreateEmpty(const ASTContext &C,
+ unsigned NumClauses);
+ static OpenACCExitDataConstruct *
+ Create(const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End, ArrayRef<const OpenACCClause *> Clauses);
+};
+// This class represents a 'host_data' construct, which has an associated
+// statement and clauses, but is otherwise pretty simple.
+class OpenACCHostDataConstruct final
+ : public OpenACCAssociatedStmtConstruct,
+ public llvm::TrailingObjects<OpenACCCombinedConstruct,
+ const OpenACCClause *> {
+ OpenACCHostDataConstruct(unsigned NumClauses)
+ : OpenACCAssociatedStmtConstruct(
+ OpenACCHostDataConstructClass, OpenACCDirectiveKind::HostData,
+ SourceLocation{}, SourceLocation{}, SourceLocation{},
+ /*AssociatedStmt=*/nullptr) {
+ std::uninitialized_value_construct(
+ getTrailingObjects<const OpenACCClause *>(),
+ getTrailingObjects<const OpenACCClause *>() + NumClauses);
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ NumClauses));
+ }
+ OpenACCHostDataConstruct(SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End,
+ ArrayRef<const OpenACCClause *> Clauses,
+ Stmt *StructuredBlock)
+ : OpenACCAssociatedStmtConstruct(OpenACCHostDataConstructClass,
+ OpenACCDirectiveKind::HostData, Start,
+ DirectiveLoc, End, StructuredBlock) {
+ std::uninitialized_copy(Clauses.begin(), Clauses.end(),
+ getTrailingObjects<const OpenACCClause *>());
+ setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(),
+ Clauses.size()));
+ }
+ void setStructuredBlock(Stmt *S) { setAssociatedStmt(S); }
+
+public:
+ static bool classof(const Stmt *T) {
+ return T->getStmtClass() == OpenACCHostDataConstructClass;
+ }
+ static OpenACCHostDataConstruct *CreateEmpty(const ASTContext &C,
+ unsigned NumClauses);
+ static OpenACCHostDataConstruct *
+ Create(const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End, ArrayRef<const OpenACCClause *> Clauses,
+ Stmt *StructuredBlock);
+ Stmt *getStructuredBlock() { return getAssociatedStmt(); }
+ const Stmt *getStructuredBlock() const {
+ return const_cast<OpenACCHostDataConstruct *>(this)->getStructuredBlock();
+ }
+};
} // namespace clang
#endif // LLVM_CLANG_AST_STMTOPENACC_H
diff --git a/clang/include/clang/AST/TextNodeDumper.h b/clang/include/clang/AST/TextNodeDumper.h
index 988b142a7672a3..e54e7e527b8a36 100644
--- a/clang/include/clang/AST/TextNodeDumper.h
+++ b/clang/include/clang/AST/TextNodeDumper.h
@@ -411,6 +411,10 @@ class TextNodeDumper
void VisitOpenACCConstructStmt(const OpenACCConstructStmt *S);
void VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *S);
void VisitOpenACCCombinedConstruct(const OpenACCCombinedConstruct *S);
+ void VisitOpenACCDataConstruct(const OpenACCDataConstruct *S);
+ void VisitOpenACCEnterDataConstruct(const OpenACCEnterDataConstruct *S);
+ void VisitOpenACCExitDataConstruct(const OpenACCExitDataConstruct *S);
+ void VisitOpenACCHostDataConstruct(const OpenACCHostDataConstruct *S);
void VisitOpenACCAsteriskSizeExpr(const OpenACCAsteriskSizeExpr *S);
void VisitEmbedExpr(const EmbedExpr *S);
void VisitAtomicExpr(const AtomicExpr *AE);
diff --git a/clang/include/clang/Basic/StmtNodes.td b/clang/include/clang/Basic/StmtNodes.td
index 89f5a76eb11312..0c3c580c218fd7 100644
--- a/clang/include/clang/Basic/StmtNodes.td
+++ b/clang/include/clang/Basic/StmtNodes.td
@@ -308,6 +308,10 @@ def OpenACCAssociatedStmtConstruct
def OpenACCComputeConstruct : StmtNode<OpenACCAssociatedStmtConstruct>;
def OpenACCLoopConstruct : StmtNode<OpenACCAssociatedStmtConstruct>;
def OpenACCCombinedConstruct : StmtNode<OpenACCAssociatedStmtConstruct>;
+def OpenACCDataConstruct : StmtNode<OpenACCAssociatedStmtConstruct>;
+def OpenACCEnterDataConstruct : StmtNode<OpenACCConstructStmt>;
+def OpenACCExitDataConstruct : StmtNode<OpenACCConstructStmt>;
+def OpenACCHostDataConstruct : StmtNode<OpenACCAssociatedStmtConstruct>;
// OpenACC Additional Expressions.
def OpenACCAsteriskSizeExpr : StmtNode<Expr>;
diff --git a/clang/include/clang/Serialization/ASTBitCodes.h b/clang/include/clang/Serialization/ASTBitCodes.h
index af0e08d800bf28..2be9ade08cac31 100644
--- a/clang/include/clang/Serialization/ASTBitCodes.h
+++ b/clang/include/clang/Serialization/ASTBitCodes.h
@@ -2019,6 +2019,10 @@ enum StmtCode {
STMT_OPENACC_LOOP_CONSTRUCT,
STMT_OPENACC_COMBINED_CONSTRUCT,
EXPR_OPENACC_ASTERISK_SIZE,
+ STMT_OPENACC_DATA_CONSTRUCT,
+ STMT_OPENACC_ENTER_DATA_CONSTRUCT,
+ STMT_OPENACC_EXIT_DATA_CONSTRUCT,
+ STMT_OPENACC_HOST_DATA_CONSTRUCT,
// HLSL Constructs
EXPR_HLSL_OUT_ARG,
diff --git a/clang/lib/AST/StmtOpenACC.cpp b/clang/lib/AST/StmtOpenACC.cpp
index 23dd57d235813a..fb73dfb3fa9dee 100644
--- a/clang/lib/AST/StmtOpenACC.cpp
+++ b/clang/lib/AST/StmtOpenACC.cpp
@@ -110,3 +110,89 @@ OpenACCCombinedConstruct *OpenACCCombinedConstruct::Create(
OpenACCCombinedConstruct(DK, BeginLoc, DirLoc, EndLoc, Clauses, Loop);
return Inst;
}
+
+OpenACCDataConstruct *OpenACCDataConstruct::CreateEmpty(const ASTContext &C,
+ unsigned NumClauses) {
+ void *Mem =
+ C.Allocate(OpenACCDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ NumClauses));
+ auto *Inst = new (Mem) OpenACCDataConstruct(NumClauses);
+ return Inst;
+}
+
+OpenACCDataConstruct *
+OpenACCDataConstruct::Create(const ASTContext &C, SourceLocation Start,
+ SourceLocation DirectiveLoc, SourceLocation End,
+ ArrayRef<const OpenACCClause *> Clauses,
+ Stmt *StructuredBlock) {
+ void *Mem =
+ C.Allocate(OpenACCDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ Clauses.size()));
+ auto *Inst = new (Mem)
+ OpenACCDataConstruct(Start, DirectiveLoc, End, Clauses, StructuredBlock);
+ return Inst;
+}
+
+OpenACCEnterDataConstruct *
+OpenACCEnterDataConstruct::CreateEmpty(const ASTContext &C,
+ unsigned NumClauses) {
+ void *Mem = C.Allocate(
+ OpenACCEnterDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ NumClauses));
+ auto *Inst = new (Mem) OpenACCEnterDataConstruct(NumClauses);
+ return Inst;
+}
+
+OpenACCEnterDataConstruct *OpenACCEnterDataConstruct::Create(
+ const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End, ArrayRef<const OpenACCClause *> Clauses) {
+ void *Mem = C.Allocate(
+ OpenACCEnterDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ Clauses.size()));
+ auto *Inst =
+ new (Mem) OpenACCEnterDataConstruct(Start, DirectiveLoc, End, Clauses);
+ return Inst;
+}
+
+OpenACCExitDataConstruct *
+OpenACCExitDataConstruct::CreateEmpty(const ASTContext &C,
+ unsigned NumClauses) {
+ void *Mem = C.Allocate(
+ OpenACCExitDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ NumClauses));
+ auto *Inst = new (Mem) OpenACCExitDataConstruct(NumClauses);
+ return Inst;
+}
+
+OpenACCExitDataConstruct *OpenACCExitDataConstruct::Create(
+ const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End, ArrayRef<const OpenACCClause *> Clauses) {
+ void *Mem = C.Allocate(
+ OpenACCExitDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ Clauses.size()));
+ auto *Inst =
+ new (Mem) OpenACCExitDataConstruct(Start, DirectiveLoc, End, Clauses);
+ return Inst;
+}
+
+OpenACCHostDataConstruct *
+OpenACCHostDataConstruct::CreateEmpty(const ASTContext &C,
+ unsigned NumClauses) {
+ void *Mem = C.Allocate(
+ OpenACCHostDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ NumClauses));
+ auto *Inst = new (Mem) OpenACCHostDataConstruct(NumClauses);
+ return Inst;
+}
+
+OpenACCHostDataConstruct *OpenACCHostDataConstruct::Create(
+ const ASTContext &C, SourceLocation Start, SourceLocation DirectiveLoc,
+ SourceLocation End, ArrayRef<const OpenACCClause *> Clauses,
+ Stmt *StructuredBlock) {
+ void *Mem = C.Allocate(
+ OpenACCHostDataConstruct::totalSizeToAlloc<const OpenACCClause *>(
+ Clauses.size()));
+ auto *Inst = new (Mem) OpenACCHostDataConstruct(Start, DirectiveLoc, End,
+ Clauses, StructuredBlock);
+ return Inst;
+}
diff --git a/clang/lib/AST/StmtPrinter.cpp b/clang/lib/AST/StmtPrinter.cpp
index 7507c9d14327a0..488419add5e79e 100644
--- a/clang/lib/AST/StmtPrinter.cpp
+++ b/clang/lib/AST/StmtPrinter.cpp
@@ -1193,6 +1193,51 @@ void StmtPrinter::VisitOpenACCCombinedConstruct(OpenACCCombinedConstruct *S) {
PrintStmt(S->getLoop());
}
+void StmtPrinter::VisitOpenACCDataConstruct(OpenACCDataConstruct *S) {
+ Indent() << "#pragma acc data";
+
+ if (!S->clauses().empty()) {
+ OS << ' ';
+ OpenACCClausePrinter Printer(OS, Policy);
+ Printer.VisitClauseList(S->clauses());
+ }
+ OS << '\n';
+
+ PrintStmt(S->getStructuredBlock());
+}
+void StmtPrinter::VisitOpenACCEnterDataConstruct(OpenACCEnterDataConstruct *S) {
+ Indent() << "#pragma acc enter data";
+
+ if (!S->clauses().empty()) {
+ OS << ' ';
+ OpenACCClausePrinter Printer(OS, Policy);
+ Printer.VisitClauseList(S->clauses());
+ }
+ OS << '\n';
+}
+void StmtPrinter::VisitOpenACCExitDataConstruct(OpenACCExitDataConstruct *S) {
+ Indent() << "#pragma acc exit data";
+
+ if (!S->clauses().empty()) {
+ OS << ' ';
+ OpenACCClausePrinter Printer(OS, Policy);
+ Printer.VisitClauseList(S->clauses());
+ }
+ OS << '\n';
+}
+void StmtPrinter::VisitOpenACCHostDataConstruct(OpenACCHostDataConstruct *S) {
+ Indent() << "#pragma acc host_data";
+
+ if (!S->clauses().empty()) {
+ OS << ' ';
+ OpenACCClausePrinter Printer(OS, Policy);
+ Printer.VisitClauseList(S->clauses());
+ }
+ OS << '\n';
+
+ PrintStmt(S->getStructuredBlock());
+}
+
//===----------------------------------------------------------------------===//
// Expr printing methods.
//===----------------------------------------------------------------------===//
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index 3dfbef1cdb712d..e9ff674097c8fa 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -2697,6 +2697,37 @@ void StmtProfiler::VisitOpenACCCombinedConstruct(
P.VisitOpenACCClauseList(S->clauses());
}
+void StmtProfiler::VisitOpenACCDataConstruct(const OpenACCDataConstruct *S) {
+ VisitStmt(S);
+
+ OpenACCClauseProfiler P{*this};
+ P.VisitOpenACCClauseList(S->clauses());
+}
+
+void StmtProfiler::VisitOpenACCEnterDataConstruct(
+ const OpenACCEnterDataConstruct *S) {
+ VisitStmt(S);
+
+ OpenACCClauseProfiler P{*this};
+ P.VisitOpenACCClauseList(S->clauses());
+}
+
+void StmtProfiler::VisitOpenACCExitDataConstruct(
+ const OpenACCExitDataConstruct *S) {
+ VisitStmt(S);
+
+ OpenACCClauseProfiler P{*this};
+ P.VisitOpenACCClauseList(S->clauses());
+}
+
+void StmtProfiler::VisitOpenACCHostDataConstruct(
+ const OpenACCHostDataConstruct *S) {
+ VisitStmt(S);
+
+ OpenACCClauseProfiler P{*this};
+ P.VisitOpenACCClauseList(S->clauses());
+}
+
void StmtProfiler::VisitHLSLOutArgExpr(const HLSLOutArgExpr *S) {
VisitStmt(S);
}
diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp
index 2552c11a395320..209ad3a5f10ac4 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -2936,6 +2936,25 @@ void TextNodeDumper::VisitOpenACCCombinedConstruct(
OS << " " << S->getDirectiveKind();
}
+void TextNodeDumper::VisitOpenACCDataConstruct(const OpenACCDataConstruct *S) {
+ OS << " " << S->getDirectiveKind();
+}
+
+void TextNodeDumper::VisitOpenACCEnterDataConstruct(
+ const OpenACCEnterDataConstruct *S) {
+ OS << " " << S->getDirectiveKind();
+}
+
+void TextNodeDumper::VisitOpenACCExitDataConstruct(
+ const OpenACCExitDataConstruct *S) {
+ OS << " " << S->getDirectiveKind();
+}
+
+void TextNodeDumper::VisitOpenACCHostDataConstruct(
+ const OpenACCHostDataConstruct *S) {
+ OS << " " << S->getDirectiveKind();
+}
+
void TextNodeDumper::VisitEmbedExpr(const EmbedExpr *S) {
AddChild("begin", [=] { OS << S->getStartingElementPos(); });
AddChild("number of elements", [=] { OS << S->getDataElementCount(); });
diff --git a/clang/lib/CodeGen/CGStmt.cpp b/clang/lib/CodeGen/CGStmt.cpp
index 698baf853507f4..6c7a594fb10c4c 100644
--- a/clang/lib/CodeGen/CGStmt.cpp
+++ b/clang/lib/CodeGen/CGStmt.cpp
@@ -458,6 +458,18 @@ void CodeGenFunction::EmitStmt(const Stmt *S, ArrayRef<const Attr *> Attrs) {
case Stmt::OpenACCCombinedConstructClass:
EmitOpenACCCombinedConstruct(cast<OpenACCCombinedConstruct>(*S));
break;
+ case Stmt::OpenACCDataConstructClass:
+ EmitOpenACCDataConstruct(cast<OpenACCDataConstruct>(*S));
+ break;
+ case Stmt::OpenACCEnterDataConstructClass:
+ EmitOpenACCEnterDataConstruct(cast<OpenACCEnterDataConstruct>(*S));
+ break;
+ case Stmt::OpenACCExitDataConstructClass:
+ EmitOpenACCExitDataConstruct(cast<OpenACCExitDataConstruct>(*S));
+ break;
+ case Stmt::OpenACCHostDataConstructClass:
+ EmitOpenACCHostDataConstruct(cast<OpenACCHostDataConstruct>(*S));
+ break;
}
}
diff --git a/clang/lib/CodeGen/CodeGenFunction.h b/clang/lib/CodeGen/CodeGenFunction.h
index eaea0d8a08ac06..092d55355a0a17 100644
--- a/clang/lib/CodeGen/CodeGenFunction.h
+++ b/clang/lib/CodeGen/CodeGenFunction.h
@@ -4094,6 +4094,30 @@ class CodeGenFunction : public CodeGenTypeCache {
EmitStmt(S.getLoop());
}
+ void EmitOpenACCDataConstruct(const OpenACCDataConstruct &S) {
+ // TODO OpenACC: Implement this. It is currently implemented as a 'no-op',
+ // simply emitting its structured block, but in the future we will implement
+ // some sort of IR.
+ EmitStmt(S.getStructuredBlock());
+ }
+
+ void EmitOpenACCEnterDataConstruct(const OpenACCEnterDataConstruct &S) {
+ // TODO OpenACC: Implement this. It is currently implemented as a 'no-op',
+ // but in the future we will implement some sort of IR.
+ }
+
+ void EmitOpenACCExitDataConstruct(const OpenACCExitDataConstruct &S) {
+ // TODO OpenACC: Implement this. It is currently implemented as a 'no-op',
+ // but in the future we will implement some sort of IR.
+ }
+
+ void EmitOpenACCHostDataConstruct(const OpenACCHostDataConstruct &S) {
+ // TODO OpenACC: Implement this. It is currently implemented as a 'no-op',
+ // simply emitting its structured block, but in the future we will implement
+ // some sort of IR.
+ EmitStmt(S.getStructuredBlock());
+ }
+
//===--------------------------------------------------------------------===//
// LValue Expression Emission
//===--------------------------------------------------------------------===//
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index bc59de3c1a0ada..8c81936b35296c 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -571,6 +571,8 @@ void SkipUntilEndOfDirective(Parser &P) {
bool doesDirectiveHaveAssociatedStmt(OpenACCDirectiveKind DirKind) {
switch (DirKind) {
default:
+ case OpenACCDirectiveKind::EnterData:
+ case OpenACCDirectiveKind::ExitData:
return false;
case OpenACCDirectiveKind::Parallel:
case OpenACCDirectiveKind::Serial:
@@ -579,6 +581,8 @@ bool doesDirectiveHaveAssociatedStmt(OpenACCDirectiveKind DirKind) {
case OpenACCDirectiveKind::SerialLoop:
case OpenACCDirectiveKind::KernelsLoop:
case OpenACCDirectiveKind::Loop:
+ case OpenACCDirectiveKind::Data:
+ case OpenACCDirectiveKind::HostData:
return true;
}
llvm_unreachable("Unhandled directive->assoc stmt");
@@ -596,6 +600,11 @@ unsigned getOpenACCScopeFlags(OpenACCDirectiveKind DirKind) {
// so that we can diagnose trying to 'break'/'continue' inside of one.
return Scope::BreakScope | Scope::ContinueScope |
Scope::OpenACCComputeConstructScope;
+ case OpenACCDirectiveKind::Data:
+ case OpenACCDirectiveKind::EnterData:
+ case OpenACCDirectiveKind::ExitData:
+ case OpenACCDirectiveKind::HostData:
+ return 0;
case OpenACCDirectiveKind::Invalid:
llvm_unreachable("Shouldn't be creating a scope for an invalid construct");
default:
@@ -1508,10 +1517,10 @@ StmtResult Parser::ParseOpenACCDirectiveStmt() {
return StmtError();
StmtResult AssocStmt;
- SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(getActions().OpenACC(),
- DirInfo.DirKind, DirInfo.DirLoc,
- {}, DirInfo.Clauses);
if (doesDirectiveHaveAssociatedStmt(DirInfo.DirKind)) {
+ SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(
+ getActions().OpenACC(), DirInfo.DirKind, DirInfo.DirLoc, {},
+ DirInfo.Clauses);
ParsingOpenACCDirectiveRAII DirScope(*this, /*Value=*/false);
ParseScope ACCScope(this, getOpenACCScopeFlags(DirInfo.DirKind));
diff --git a/clang/lib/Sema/SemaExceptionSpec.cpp b/clang/lib/Sema/SemaExceptionSpec.cpp
index 6a9f43d6f5215e..2be6af293ed543 100644
--- a/clang/lib/Sema/SemaExceptionSpec.cpp
+++ b/clang/lib/Sema/SemaExceptionSpec.cpp
@@ -1396,6 +1396,8 @@ CanThrowResult Sema::canThrow(const Stmt *S) {
case Expr::ConceptSpecializationExprClass:
case Expr::RequiresExprClass:
case Expr::HLSLOutArgExprClass:
+ case Stmt::OpenACCEnterDataConstructClass:
+ case Stmt::OpenACCExitDataConstructClass:
// These expressions can never throw.
return CT_Cannot;
@@ -1407,6 +1409,8 @@ CanThrowResult Sema::canThrow(const Stmt *S) {
case Stmt::OpenACCComputeConstructClass:
case Stmt::OpenACCLoopConstructClass:
case Stmt::OpenACCCombinedConstructClass:
+ case Stmt::OpenACCDataConstructClass:
+ case Stmt::OpenACCHostDataConstructClass:
case Stmt::AttributedStmtClass:
case Stmt::BreakStmtClass:
case Stmt::CapturedStmtClass:
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 62c3e778ab178d..5575dd730e5596 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -37,6 +37,10 @@ bool diagnoseConstructAppertainment(SemaOpenACC &S, OpenACCDirectiveKind K,
case OpenACCDirectiveKind::Serial:
case OpenACCDirectiveKind::Kernels:
case OpenACCDirectiveKind::Loop:
+ case OpenACCDirectiveKind::Data:
+ case OpenACCDirectiveKind::EnterData:
+ case OpenACCDirectiveKind::ExitData:
+ case OpenACCDirectiveKind::HostData:
if (!IsStmt)
return S.Diag(StartLoc, diag::err_acc_construct_appertainment) << K;
break;
@@ -1760,6 +1764,31 @@ void CollectActiveReductionClauses(
}
}
+// Depth needs to be preserved for all associated statements that aren't
+// supposed to modify the compute/combined/loop construct information.
+bool PreserveLoopRAIIDepthInAssociatedStmtRAII(OpenACCDirectiveKind DK) {
+ switch (DK) {
+ case OpenACCDirectiveKind::Parallel:
+ case OpenACCDirectiveKind::ParallelLoop:
+ case OpenACCDirectiveKind::Serial:
+ case OpenACCDirectiveKind::SerialLoop:
+ case OpenACCDirectiveKind::Kernels:
+ case OpenACCDirectiveKind::KernelsLoop:
+ case OpenACCDirectiveKind::Loop:
+ return false;
+ case OpenACCDirectiveKind::Data:
+ case OpenACCDirectiveKind::HostData:
+ return true;
+ case OpenACCDirectiveKind::EnterData:
+ case OpenACCDirectiveKind::ExitData:
+ llvm_unreachable("Doesn't have an associated stmt");
+ default:
+ case OpenACCDirectiveKind::Invalid:
+ llvm_unreachable("Unhandled directive kind?");
+ }
+ llvm_unreachable("Unhandled directive kind?");
+}
+
} // namespace
SemaOpenACC::SemaOpenACC(Sema &S) : SemaBase(S) {}
@@ -1774,7 +1803,7 @@ SemaOpenACC::AssociatedStmtRAII::AssociatedStmtRAII(
OldLoopVectorClauseLoc(S.LoopVectorClauseLoc),
OldLoopWithoutSeqInfo(S.LoopWithoutSeqInfo),
ActiveReductionClauses(S.ActiveReductionClauses),
- LoopRAII(SemaRef, /*PreserveDepth=*/false) {
+ LoopRAII(SemaRef, PreserveLoopRAIIDepthInAssociatedStmtRAII(DirKind)) {
// Compute constructs end up taking their 'loop'.
if (DirKind == OpenACCDirectiveKind::Parallel ||
@@ -1950,24 +1979,23 @@ void SemaOpenACC::AssociatedStmtRAII::SetTileInfoBeforeAssociatedStmt(
}
SemaOpenACC::AssociatedStmtRAII::~AssociatedStmtRAII() {
- SemaRef.ActiveComputeConstructInfo = OldActiveComputeConstructInfo;
- SemaRef.LoopGangClauseOnKernel = OldLoopGangClauseOnKernel;
- SemaRef.LoopWorkerClauseLoc = OldLoopWorkerClauseLoc;
- SemaRef.LoopVectorClauseLoc = OldLoopVectorClauseLoc;
- SemaRef.LoopWithoutSeqInfo = OldLoopWithoutSeqInfo;
- SemaRef.ActiveReductionClauses.swap(ActiveReductionClauses);
-
if (DirKind == OpenACCDirectiveKind::Parallel ||
DirKind == OpenACCDirectiveKind::Serial ||
DirKind == OpenACCDirectiveKind::Kernels ||
+ DirKind == OpenACCDirectiveKind::Loop ||
DirKind == OpenACCDirectiveKind::ParallelLoop ||
DirKind == OpenACCDirectiveKind::SerialLoop ||
DirKind == OpenACCDirectiveKind::KernelsLoop) {
- // Nothing really to do here, the restorations above should be enough for
- // now.
- } else if (DirKind == OpenACCDirectiveKind::Loop) {
- // Nothing really to do here, the LoopInConstruct should handle restorations
- // correctly.
+ SemaRef.ActiveComputeConstructInfo = OldActiveComputeConstructInfo;
+ SemaRef.LoopGangClauseOnKernel = OldLoopGangClauseOnKernel;
+ SemaRef.LoopWorkerClauseLoc = OldLoopWorkerClauseLoc;
+ SemaRef.LoopVectorClauseLoc = OldLoopVectorClauseLoc;
+ SemaRef.LoopWithoutSeqInfo = OldLoopWithoutSeqInfo;
+ SemaRef.ActiveReductionClauses.swap(ActiveReductionClauses);
+ } else if (DirKind == OpenACCDirectiveKind::Data ||
+ DirKind == OpenACCDirectiveKind::HostData) {
+ // Intentionally doesn't reset the Loop, Compute Construct, or reduction
+ // effects.
}
}
@@ -2175,6 +2203,10 @@ void SemaOpenACC::ActOnConstruct(OpenACCDirectiveKind K,
case OpenACCDirectiveKind::SerialLoop:
case OpenACCDirectiveKind::KernelsLoop:
case OpenACCDirectiveKind::Loop:
+ case OpenACCDirectiveKind::Data:
+ case OpenACCDirectiveKind::EnterData:
+ case OpenACCDirectiveKind::ExitData:
+ case OpenACCDirectiveKind::HostData:
// Nothing to do here, there is no real legalization that needs to happen
// here as these constructs do not take any arguments.
break;
@@ -3441,6 +3473,24 @@ StmtResult SemaOpenACC::ActOnEndStmtDirective(OpenACCDirectiveKind K,
getASTContext(), ActiveComputeConstructInfo.Kind, StartLoc, DirLoc,
EndLoc, Clauses, AssocStmt.isUsable() ? AssocStmt.get() : nullptr);
}
+ case OpenACCDirectiveKind::Data: {
+ return OpenACCDataConstruct::Create(
+ getASTContext(), StartLoc, DirLoc, EndLoc, Clauses,
+ AssocStmt.isUsable() ? AssocStmt.get() : nullptr);
+ }
+ case OpenACCDirectiveKind::EnterData: {
+ return OpenACCEnterDataConstruct::Create(getASTContext(), StartLoc, DirLoc,
+ EndLoc, Clauses);
+ }
+ case OpenACCDirectiveKind::ExitData: {
+ return OpenACCExitDataConstruct::Create(getASTContext(), StartLoc, DirLoc,
+ EndLoc, Clauses);
+ }
+ case OpenACCDirectiveKind::HostData: {
+ return OpenACCHostDataConstruct::Create(
+ getASTContext(), StartLoc, DirLoc, EndLoc, Clauses,
+ AssocStmt.isUsable() ? AssocStmt.get() : nullptr);
+ }
}
llvm_unreachable("Unhandled case in directive handling?");
}
@@ -3451,9 +3501,15 @@ StmtResult SemaOpenACC::ActOnAssociatedStmt(
switch (K) {
default:
llvm_unreachable("Unimplemented associated statement application");
+ case OpenACCDirectiveKind::EnterData:
+ case OpenACCDirectiveKind::ExitData:
+ llvm_unreachable(
+ "these don't have associated statements, so shouldn't get here");
case OpenACCDirectiveKind::Parallel:
case OpenACCDirectiveKind::Serial:
case OpenACCDirectiveKind::Kernels:
+ case OpenACCDirectiveKind::Data:
+ case OpenACCDirectiveKind::HostData:
// There really isn't any checking here that could happen. As long as we
// have a statement to associate, this should be fine.
// OpenACC 3.3 Section 6:
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index 02d2fc018e3c35..f2dbf4086a13da 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -4110,6 +4110,42 @@ class TreeTransform {
EndLoc, Clauses, Loop);
}
+ StmtResult RebuildOpenACCDataConstruct(SourceLocation BeginLoc,
+ SourceLocation DirLoc,
+ SourceLocation EndLoc,
+ ArrayRef<OpenACCClause *> Clauses,
+ StmtResult StrBlock) {
+ return getSema().OpenACC().ActOnEndStmtDirective(OpenACCDirectiveKind::Data,
+ BeginLoc, DirLoc, EndLoc,
+ Clauses, StrBlock);
+ }
+
+ StmtResult
+ RebuildOpenACCEnterDataConstruct(SourceLocation BeginLoc,
+ SourceLocation DirLoc, SourceLocation EndLoc,
+ ArrayRef<OpenACCClause *> Clauses) {
+ return getSema().OpenACC().ActOnEndStmtDirective(
+ OpenACCDirectiveKind::EnterData, BeginLoc, DirLoc, EndLoc, Clauses, {});
+ }
+
+ StmtResult
+ RebuildOpenACCExitDataConstruct(SourceLocation BeginLoc,
+ SourceLocation DirLoc, SourceLocation EndLoc,
+ ArrayRef<OpenACCClause *> Clauses) {
+ return getSema().OpenACC().ActOnEndStmtDirective(
+ OpenACCDirectiveKind::ExitData, BeginLoc, DirLoc, EndLoc, Clauses, {});
+ }
+
+ StmtResult RebuildOpenACCHostDataConstruct(SourceLocation BeginLoc,
+ SourceLocation DirLoc,
+ SourceLocation EndLoc,
+ ArrayRef<OpenACCClause *> Clauses,
+ StmtResult StrBlock) {
+ return getSema().OpenACC().ActOnEndStmtDirective(
+ OpenACCDirectiveKind::HostData, BeginLoc, DirLoc, EndLoc, Clauses,
+ StrBlock);
+ }
+
ExprResult RebuildOpenACCAsteriskSizeExpr(SourceLocation AsteriskLoc) {
return getSema().OpenACC().ActOnOpenACCAsteriskSizeExpr(AsteriskLoc);
}
@@ -12153,6 +12189,88 @@ StmtResult TreeTransform<Derived>::TransformOpenACCCombinedConstruct(
C->getEndLoc(), TransformedClauses, Loop);
}
+template <typename Derived>
+StmtResult
+TreeTransform<Derived>::TransformOpenACCDataConstruct(OpenACCDataConstruct *C) {
+ getSema().OpenACC().ActOnConstruct(C->getDirectiveKind(), C->getBeginLoc());
+
+ llvm::SmallVector<OpenACCClause *> TransformedClauses =
+ getDerived().TransformOpenACCClauseList(C->getDirectiveKind(),
+ C->clauses());
+ if (getSema().OpenACC().ActOnStartStmtDirective(C->getDirectiveKind(),
+ C->getBeginLoc()))
+ return StmtError();
+
+ SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(
+ getSema().OpenACC(), C->getDirectiveKind(), C->getDirectiveLoc(),
+ C->clauses(), TransformedClauses);
+ StmtResult StrBlock = getDerived().TransformStmt(C->getStructuredBlock());
+ StrBlock = getSema().OpenACC().ActOnAssociatedStmt(
+ C->getBeginLoc(), C->getDirectiveKind(), TransformedClauses, StrBlock);
+
+ return getDerived().RebuildOpenACCDataConstruct(
+ C->getBeginLoc(), C->getDirectiveLoc(), C->getEndLoc(),
+ TransformedClauses, StrBlock);
+}
+
+template <typename Derived>
+StmtResult TreeTransform<Derived>::TransformOpenACCEnterDataConstruct(
+ OpenACCEnterDataConstruct *C) {
+ getSema().OpenACC().ActOnConstruct(C->getDirectiveKind(), C->getBeginLoc());
+
+ llvm::SmallVector<OpenACCClause *> TransformedClauses =
+ getDerived().TransformOpenACCClauseList(C->getDirectiveKind(),
+ C->clauses());
+ if (getSema().OpenACC().ActOnStartStmtDirective(C->getDirectiveKind(),
+ C->getBeginLoc()))
+ return StmtError();
+
+ return getDerived().RebuildOpenACCEnterDataConstruct(
+ C->getBeginLoc(), C->getDirectiveLoc(), C->getEndLoc(),
+ TransformedClauses);
+}
+
+template <typename Derived>
+StmtResult TreeTransform<Derived>::TransformOpenACCExitDataConstruct(
+ OpenACCExitDataConstruct *C) {
+ getSema().OpenACC().ActOnConstruct(C->getDirectiveKind(), C->getBeginLoc());
+
+ llvm::SmallVector<OpenACCClause *> TransformedClauses =
+ getDerived().TransformOpenACCClauseList(C->getDirectiveKind(),
+ C->clauses());
+ if (getSema().OpenACC().ActOnStartStmtDirective(C->getDirectiveKind(),
+ C->getBeginLoc()))
+ return StmtError();
+
+ return getDerived().RebuildOpenACCExitDataConstruct(
+ C->getBeginLoc(), C->getDirectiveLoc(), C->getEndLoc(),
+ TransformedClauses);
+}
+
+template <typename Derived>
+StmtResult TreeTransform<Derived>::TransformOpenACCHostDataConstruct(
+ OpenACCHostDataConstruct *C) {
+ getSema().OpenACC().ActOnConstruct(C->getDirectiveKind(), C->getBeginLoc());
+
+ llvm::SmallVector<OpenACCClause *> TransformedClauses =
+ getDerived().TransformOpenACCClauseList(C->getDirectiveKind(),
+ C->clauses());
+ if (getSema().OpenACC().ActOnStartStmtDirective(C->getDirectiveKind(),
+ C->getBeginLoc()))
+ return StmtError();
+
+ SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(
+ getSema().OpenACC(), C->getDirectiveKind(), C->getDirectiveLoc(),
+ C->clauses(), TransformedClauses);
+ StmtResult StrBlock = getDerived().TransformStmt(C->getStructuredBlock());
+ StrBlock = getSema().OpenACC().ActOnAssociatedStmt(
+ C->getBeginLoc(), C->getDirectiveKind(), TransformedClauses, StrBlock);
+
+ return getDerived().RebuildOpenACCHostDataConstruct(
+ C->getBeginLoc(), C->getDirectiveLoc(), C->getEndLoc(),
+ TransformedClauses, StrBlock);
+}
+
template <typename Derived>
ExprResult TreeTransform<Derived>::TransformOpenACCAsteriskSizeExpr(
OpenACCAsteriskSizeExpr *E) {
diff --git a/clang/lib/Serialization/ASTReaderStmt.cpp b/clang/lib/Serialization/ASTReaderStmt.cpp
index 9f4877b19d8705..21ad6c5a9faa33 100644
--- a/clang/lib/Serialization/ASTReaderStmt.cpp
+++ b/clang/lib/Serialization/ASTReaderStmt.cpp
@@ -2849,6 +2849,27 @@ void ASTStmtReader::VisitOpenACCCombinedConstruct(OpenACCCombinedConstruct *S) {
VisitOpenACCAssociatedStmtConstruct(S);
}
+void ASTStmtReader::VisitOpenACCDataConstruct(OpenACCDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCAssociatedStmtConstruct(S);
+}
+
+void ASTStmtReader::VisitOpenACCEnterDataConstruct(
+ OpenACCEnterDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCConstructStmt(S);
+}
+
+void ASTStmtReader::VisitOpenACCExitDataConstruct(OpenACCExitDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCConstructStmt(S);
+}
+
+void ASTStmtReader::VisitOpenACCHostDataConstruct(OpenACCHostDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCAssociatedStmtConstruct(S);
+}
+
//===----------------------------------------------------------------------===//
// HLSL Constructs/Directives.
//===----------------------------------------------------------------------===//
@@ -4324,6 +4345,26 @@ Stmt *ASTReader::ReadStmtFromStream(ModuleFile &F) {
S = OpenACCCombinedConstruct::CreateEmpty(Context, NumClauses);
break;
}
+ case STMT_OPENACC_DATA_CONSTRUCT: {
+ unsigned NumClauses = Record[ASTStmtReader::NumStmtFields];
+ S = OpenACCDataConstruct::CreateEmpty(Context, NumClauses);
+ break;
+ }
+ case STMT_OPENACC_ENTER_DATA_CONSTRUCT: {
+ unsigned NumClauses = Record[ASTStmtReader::NumStmtFields];
+ S = OpenACCEnterDataConstruct::CreateEmpty(Context, NumClauses);
+ break;
+ }
+ case STMT_OPENACC_EXIT_DATA_CONSTRUCT: {
+ unsigned NumClauses = Record[ASTStmtReader::NumStmtFields];
+ S = OpenACCExitDataConstruct::CreateEmpty(Context, NumClauses);
+ break;
+ }
+ case STMT_OPENACC_HOST_DATA_CONSTRUCT: {
+ unsigned NumClauses = Record[ASTStmtReader::NumStmtFields];
+ S = OpenACCHostDataConstruct::CreateEmpty(Context, NumClauses);
+ break;
+ }
case EXPR_REQUIRES: {
unsigned numLocalParameters = Record[ASTStmtReader::NumExprFields];
unsigned numRequirement = Record[ASTStmtReader::NumExprFields + 1];
diff --git a/clang/lib/Serialization/ASTWriterStmt.cpp b/clang/lib/Serialization/ASTWriterStmt.cpp
index 603aa5707ce9be..e55cbe1f6ecce6 100644
--- a/clang/lib/Serialization/ASTWriterStmt.cpp
+++ b/clang/lib/Serialization/ASTWriterStmt.cpp
@@ -2926,6 +2926,31 @@ void ASTStmtWriter::VisitOpenACCCombinedConstruct(OpenACCCombinedConstruct *S) {
Code = serialization::STMT_OPENACC_COMBINED_CONSTRUCT;
}
+void ASTStmtWriter::VisitOpenACCDataConstruct(OpenACCDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCAssociatedStmtConstruct(S);
+ Code = serialization::STMT_OPENACC_DATA_CONSTRUCT;
+}
+
+void ASTStmtWriter::VisitOpenACCEnterDataConstruct(
+ OpenACCEnterDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCConstructStmt(S);
+ Code = serialization::STMT_OPENACC_ENTER_DATA_CONSTRUCT;
+}
+
+void ASTStmtWriter::VisitOpenACCExitDataConstruct(OpenACCExitDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCConstructStmt(S);
+ Code = serialization::STMT_OPENACC_EXIT_DATA_CONSTRUCT;
+}
+
+void ASTStmtWriter::VisitOpenACCHostDataConstruct(OpenACCHostDataConstruct *S) {
+ VisitStmt(S);
+ VisitOpenACCAssociatedStmtConstruct(S);
+ Code = serialization::STMT_OPENACC_HOST_DATA_CONSTRUCT;
+}
+
//===----------------------------------------------------------------------===//
// HLSL Constructs/Directives.
//===----------------------------------------------------------------------===//
diff --git a/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp b/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp
index b46cd9fe86fc11..ae43c59511bfa7 100644
--- a/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp
+++ b/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp
@@ -1825,6 +1825,10 @@ void ExprEngine::Visit(const Stmt *S, ExplodedNode *Pred,
case Stmt::OpenACCComputeConstructClass:
case Stmt::OpenACCLoopConstructClass:
case Stmt::OpenACCCombinedConstructClass:
+ case Stmt::OpenACCDataConstructClass:
+ case Stmt::OpenACCEnterDataConstructClass:
+ case Stmt::OpenACCExitDataConstructClass:
+ case Stmt::OpenACCHostDataConstructClass:
case Stmt::OMPUnrollDirectiveClass:
case Stmt::OMPMetaDirectiveClass:
case Stmt::HLSLOutArgExprClass: {
diff --git a/clang/test/AST/ast-print-openacc-data-construct.cpp b/clang/test/AST/ast-print-openacc-data-construct.cpp
new file mode 100644
index 00000000000000..fc15add15c6b89
--- /dev/null
+++ b/clang/test/AST/ast-print-openacc-data-construct.cpp
@@ -0,0 +1,25 @@
+// RUN: %clang_cc1 -fopenacc -Wno-openacc-deprecated-clause-alias -Wno-source-uses-openacc -ast-print %s -o - | FileCheck %s
+
+void foo() {
+ int Var;
+ // TODO OpenACC: These are only legal if they have one of a list of clauses on
+ // them, so the 'check' lines should start to include those once we implement
+ // them. For now, they don't emit those because they are 'not implemented'.
+
+// CHECK: #pragma acc data
+// CHECK-NOT: default(none)
+#pragma acc data default(none)
+ ;
+// CHECK: #pragma acc enter data
+// CHECK-NOT: copyin(Var)
+#pragma acc enter data copyin(Var)
+ ;
+// CHECK: #pragma acc exit data
+// CHECK-NOT: copyout(Var)
+#pragma acc exit data copyout(Var)
+ ;
+// CHECK: #pragma acc host_data
+// CHECK-NOT: use_device(Var)
+#pragma acc host_data use_device(Var)
+ ;
+}
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index 3741ed099cf5c2..e2f0a753dd3780 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -4,37 +4,30 @@
void func() {
- // expected-warning at +2{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
#pragma acc enter data finalize
- // expected-warning at +3{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
// expected-warning at +2{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
#pragma acc enter data finalize finalize
- // expected-warning at +3{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
- // expected-error at +2{{invalid OpenACC clause 'invalid'}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}}
+ // expected-warning at +2{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
+ // expected-error at +1{{invalid OpenACC clause 'invalid'}}
#pragma acc enter data finalize invalid
- // expected-warning at +3{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
- // expected-error at +2{{invalid OpenACC clause 'invalid'}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}}
+ // expected-warning at +2{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
+ // expected-error at +1{{invalid OpenACC clause 'invalid'}}
#pragma acc enter data finalize invalid invalid finalize
- // expected-warning at +3{{OpenACC clause 'wait' not yet implemented, clause ignored}}
- // expected-warning at +2{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}}
+ // expected-warning at +2{{OpenACC clause 'wait' not yet implemented, clause ignored}}
+ // expected-warning at +1{{OpenACC clause 'finalize' not yet implemented, clause ignored}}
#pragma acc enter data wait finalize
- // expected-warning at +2{{OpenACC clause 'if_present' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'if_present' not yet implemented, clause ignored}}
#pragma acc host_data if_present
- // expected-warning at +3{{OpenACC clause 'if_present' not yet implemented, clause ignored}}
// expected-warning at +2{{OpenACC clause 'if_present' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'if_present' not yet implemented, clause ignored}}
#pragma acc host_data if_present, if_present
// expected-error at +4{{OpenACC clause 'independent' on 'loop' construct conflicts with previous data dependence clause}}
@@ -528,27 +521,23 @@ void VarListClauses() {
#pragma acc serial firstprivate(s.array[s.value : 5], s.value), self
for(int i = 0; i < 5;++i) {}
- // expected-warning at +4{{OpenACC construct 'exit data' not yet implemented}}
// expected-error at +3{{expected ','}}
// expected-warning at +2{{OpenACC clause 'delete' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'async' not yet implemented, clause ignored}}
#pragma acc exit data delete(s.array[s.value] s.array[s.value :5] ) async
for(int i = 0; i < 5;++i) {}
- // expected-warning at +3{{OpenACC construct 'exit data' not yet implemented}}
// expected-warning at +2{{OpenACC clause 'delete' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'async' not yet implemented, clause ignored}}
#pragma acc exit data delete(s.array[s.value : 5], s.value),async
for(int i = 0; i < 5;++i) {}
- // expected-warning at +4{{OpenACC construct 'exit data' not yet implemented}}
// expected-error at +3{{expected ','}}
// expected-warning at +2{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'async' not yet implemented, clause ignored}}
#pragma acc exit data use_device(s.array[s.value] s.array[s.value :5] ),async
for(int i = 0; i < 5;++i) {}
- // expected-warning at +3{{OpenACC construct 'exit data' not yet implemented}}
// expected-warning at +2{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'async' not yet implemented, clause ignored}}
#pragma acc exit data use_device(s.array[s.value : 5], s.value), async
diff --git a/clang/test/ParserOpenACC/parse-clauses.cpp b/clang/test/ParserOpenACC/parse-clauses.cpp
index 4dc966ea9879f9..1781a279407543 100644
--- a/clang/test/ParserOpenACC/parse-clauses.cpp
+++ b/clang/test/ParserOpenACC/parse-clauses.cpp
@@ -35,7 +35,6 @@ void templ() {
#pragma acc parallel async
for(;;){}
- // expected-warning at +2{{OpenACC construct 'exit data' not yet implemented}}
// expected-warning at +1{{OpenACC clause 'delete' not yet implemented, clause ignored}}
#pragma acc exit data delete(I)
;
diff --git a/clang/test/ParserOpenACC/parse-constructs.c b/clang/test/ParserOpenACC/parse-constructs.c
index 27b9a6993fd3e8..d3b1ccb48c0349 100644
--- a/clang/test/ParserOpenACC/parse-constructs.c
+++ b/clang/test/ParserOpenACC/parse-constructs.c
@@ -54,16 +54,13 @@ void func() {
// expected-error at +1{{invalid OpenACC clause 'clause'}}
#pragma acc kernels clause list
for(;;){}
- // expected-error at +2{{invalid OpenACC clause 'clause'}}
- // expected-warning at +1{{OpenACC construct 'data' not yet implemented, pragma ignored}}
+ // expected-error at +1{{invalid OpenACC clause 'clause'}}
#pragma acc data clause list
for(;;){}
- // expected-error at +2{{invalid OpenACC clause 'clause'}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}}
+ // expected-error at +1{{invalid OpenACC clause 'clause'}}
#pragma acc enter data clause list
for(;;){}
- // expected-error at +2{{invalid OpenACC clause 'clause'}}
- // expected-warning at +1{{OpenACC construct 'exit data' not yet implemented, pragma ignored}}
+ // expected-error at +1{{invalid OpenACC clause 'clause'}}
#pragma acc exit data clause list
for(;;){}
// expected-error at +1{{invalid OpenACC directive 'enter invalid'}}
@@ -78,8 +75,7 @@ void func() {
// expected-error at +1{{expected identifier}}
#pragma acc exit }
for(;;){}
- // expected-error at +2{{invalid OpenACC clause 'clause'}}
- // expected-warning at +1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}}
+ // expected-error at +1{{invalid OpenACC clause 'clause'}}
#pragma acc host_data clause list
for(;;){}
// expected-error at +1{{invalid OpenACC clause 'clause'}}
diff --git a/clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp b/clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp
index c7db9669a9879b..31078ea7a0de9e 100644
--- a/clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp
+++ b/clang/test/SemaOpenACC/combined-construct-collapse-clause.cpp
@@ -214,14 +214,15 @@ void no_other_directives() {
#pragma acc serial loop collapse(2)
for(unsigned i = 0; i < 5; ++i) {
for(unsigned j = 0; j < 5; ++j) {
-#pragma acc data // expected-warning{{OpenACC construct 'data' not yet implemented}}
+#pragma acc data
+ ;
}
}
// expected-note at +1{{active 'collapse' clause defined here}}
#pragma acc kernels loop collapse(2)
for(unsigned i = 0; i < 5; ++i) {
// expected-error at +1{{OpenACC 'data' construct cannot appear in intervening code of a 'kernels loop' with a 'collapse' clause}}
-#pragma acc data // expected-warning{{OpenACC construct 'data' not yet implemented}}
+#pragma acc data
for(unsigned j = 0; j < 5; ++j) {
}
}
diff --git a/clang/test/SemaOpenACC/combined-construct-default-clause.c b/clang/test/SemaOpenACC/combined-construct-default-clause.c
index a9c90240cb1222..7e384ccfc17a07 100644
--- a/clang/test/SemaOpenACC/combined-construct-default-clause.c
+++ b/clang/test/SemaOpenACC/combined-construct-default-clause.c
@@ -28,7 +28,6 @@ void SingleOnly() {
#pragma acc kernels loop default(none)
for(int i = 0; i < 5; ++i);
- // expected-warning at +2{{OpenACC construct 'data' not yet implemented}}
// expected-warning at +1{{OpenACC clause 'default' not yet implemented}}
#pragma acc data default(none)
while(0);
diff --git a/clang/test/SemaOpenACC/combined-construct-if-clause.c b/clang/test/SemaOpenACC/combined-construct-if-clause.c
index 563f1cd25377bd..c0069c4ee9ef44 100644
--- a/clang/test/SemaOpenACC/combined-construct-if-clause.c
+++ b/clang/test/SemaOpenACC/combined-construct-if-clause.c
@@ -43,7 +43,6 @@ void BoolExpr(int *I, float *F) {
#pragma acc kernels loop if (*I < *F)
for (unsigned i = 0; i < 5; ++i);
- // expected-warning at +2{{OpenACC construct 'data' not yet implemented}}
// expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
#pragma acc data if (*I < *F)
for (unsigned i = 0; i < 5; ++i);
diff --git a/clang/test/SemaOpenACC/compute-construct-default-clause.c b/clang/test/SemaOpenACC/compute-construct-default-clause.c
index 70e29f3e8ac051..4aef2cbd7aec4c 100644
--- a/clang/test/SemaOpenACC/compute-construct-default-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-default-clause.c
@@ -28,7 +28,6 @@ void SingleOnly() {
#pragma acc kernels default(none)
for(int i = 0; i < 5; ++i);
- // expected-warning at +2{{OpenACC construct 'data' not yet implemented}}
// expected-warning at +1{{OpenACC clause 'default' not yet implemented}}
#pragma acc data default(none)
while(0);
diff --git a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
index 0ae972d2a99ff4..2f4a037529b500 100644
--- a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
@@ -34,11 +34,9 @@ void uses() {
#pragma acc kernels dtype(MACRO)
while(1);
- // expected-error at +2{{OpenACC 'device_type' clause is not valid on 'enter data' directive}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device_type' clause is not valid on 'enter data' directive}}
#pragma acc enter data device_type(I)
- // expected-error at +2{{OpenACC 'dtype' clause is not valid on 'enter data' directive}}
- // expected-warning at +1{{OpenACC construct 'enter data' not yet implemented}}
+ // expected-error at +1{{OpenACC 'dtype' clause is not valid on 'enter data' directive}}
#pragma acc enter data dtype(I)
diff --git a/clang/test/SemaOpenACC/compute-construct-if-clause.c b/clang/test/SemaOpenACC/compute-construct-if-clause.c
index 7cdc35275acce0..1336bf2549f3ca 100644
--- a/clang/test/SemaOpenACC/compute-construct-if-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-if-clause.c
@@ -43,7 +43,6 @@ void BoolExpr(int *I, float *F) {
#pragma acc kernels if (*I < *F)
while(0);
- // expected-warning at +2{{OpenACC construct 'data' not yet implemented}}
// expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
#pragma acc data if (*I < *F)
while(0);
diff --git a/clang/test/SemaOpenACC/data-construct-ast.cpp b/clang/test/SemaOpenACC/data-construct-ast.cpp
new file mode 100644
index 00000000000000..9a7fe2cb793a73
--- /dev/null
+++ b/clang/test/SemaOpenACC/data-construct-ast.cpp
@@ -0,0 +1,88 @@
+// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
+
+// Test this with PCH.
+// RUN: %clang_cc1 %s -fopenacc -emit-pch -o %t %s
+// RUN: %clang_cc1 %s -fopenacc -include-pch %t -ast-dump-all | FileCheck %s
+
+#ifndef PCH_HELPER
+#define PCH_HELPER
+
+void NormalFunc() {
+ // CHECK-LABEL: NormalFunc
+ // CHECK-NEXT: CompoundStmt
+
+ int Var;
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl
+
+ // TODO OpenACC: these constructs require the clauses to be legal, but we
+ // don't have the clauses implemented yet. As we implement them, they needed
+ // to be added to the 'check' lines.
+
+#pragma acc data default(none)
+ while (Var);
+ // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+ // CHECK-NEXT: WhileStmt
+ // CHECK: NullStmt
+#pragma acc enter data copyin(Var)
+ // CHECK-NEXT: OpenACCEnterDataConstruct{{.*}} enter data
+#pragma acc exit data copyout(Var)
+ // CHECK-NEXT: OpenACCExitDataConstruct{{.*}} exit data
+#pragma acc host_data use_device(Var)
+ while (Var);
+ // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data
+ // CHECK-NEXT: WhileStmt
+ // CHECK: NullStmt
+}
+
+template<typename T>
+void TemplFunc() {
+ // CHECK-LABEL: FunctionTemplateDecl {{.*}}TemplFunc
+ // CHECK-NEXT: TemplateTypeParmDecl
+ // CHECK-NEXT: FunctionDecl{{.*}}TemplFunc
+ // CHECK-NEXT: CompoundStmt
+
+ T Var;
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl
+
+#pragma acc data default(none)
+ while (Var);
+ // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+ // CHECK-NEXT: WhileStmt
+ // CHECK: NullStmt
+#pragma acc enter data copyin(Var)
+ // CHECK-NEXT: OpenACCEnterDataConstruct{{.*}} enter data
+#pragma acc exit data copyout(Var)
+ // CHECK-NEXT: OpenACCExitDataConstruct{{.*}} exit data
+#pragma acc host_data use_device(Var)
+ while (Var);
+ // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data
+ // CHECK-NEXT: WhileStmt
+ // CHECK: NullStmt
+
+ // Instantiation:
+ // CHECK-NEXT: FunctionDecl{{.*}} TemplFunc 'void ()' implicit_instantiation
+ // CHECK-NEXT: TemplateArgument type 'int'
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl
+
+ // CHECK-NEXT: OpenACCDataConstruct{{.*}}data
+ // CHECK-NEXT: WhileStmt
+ // CHECK: NullStmt
+
+ // CHECK-NEXT: OpenACCEnterDataConstruct{{.*}} enter data
+
+ // CHECK-NEXT: OpenACCExitDataConstruct{{.*}} exit data
+
+ // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data
+ // CHECK-NEXT: WhileStmt
+ // CHECK: NullStmt
+}
+void use() {
+ TemplFunc<int>();
+}
+#endif
diff --git a/clang/test/SemaOpenACC/data-construct.cpp b/clang/test/SemaOpenACC/data-construct.cpp
new file mode 100644
index 00000000000000..0c1959dc427248
--- /dev/null
+++ b/clang/test/SemaOpenACC/data-construct.cpp
@@ -0,0 +1,237 @@
+// RUN: %clang_cc1 %s -fopenacc -verify -Wno-empty-body -Wno-unused-value
+
+void HasStmt() {
+ {
+ // expected-error at +2{{expected statement}}
+#pragma acc data
+ }
+ {
+ // expected-error at +2{{expected statement}}
+#pragma acc host_data
+ }
+ // Don't have statements, so this is fine.
+ {
+#pragma acc enter data
+ }
+ {
+#pragma acc exit data
+ }
+}
+
+void AtLeastOneOf() {
+ int Var;
+// Data
+ // expected-warning at +1{{OpenACC clause 'copy' not yet implemented}}
+#pragma acc data copy(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'copyin' not yet implemented}}
+#pragma acc data copyin(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'copyout' not yet implemented}}
+#pragma acc data copyout(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'create' not yet implemented}}
+#pragma acc data create(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'no_create' not yet implemented}}
+#pragma acc data no_create(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
+#pragma acc data present(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'deviceptr' not yet implemented}}
+#pragma acc data deviceptr(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'attach' not yet implemented}}
+#pragma acc data attach(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'default' not yet implemented}}
+#pragma acc data default(none)
+ ;
+
+ // OpenACC TODO: The following 'data' directives should diagnose, since they
+ // don't have at least one of the above clauses.
+
+ // expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
+#pragma acc data if(Var)
+ ;
+
+ // expected-warning at +1{{OpenACC clause 'async' not yet implemented}}
+#pragma acc data async
+ ;
+
+ // expected-warning at +1{{OpenACC clause 'wait' not yet implemented}}
+#pragma acc data wait
+ ;
+
+ // expected-warning at +1{{OpenACC clause 'device_type' not yet implemented}}
+#pragma acc data device_type(*)
+ ;
+#pragma acc data
+ ;
+
+ // Enter Data
+ // expected-warning at +1{{OpenACC clause 'copyin' not yet implemented}}
+#pragma acc enter data copyin(Var)
+ // expected-warning at +1{{OpenACC clause 'create' not yet implemented}}
+#pragma acc enter data create(Var)
+ // expected-warning at +1{{OpenACC clause 'attach' not yet implemented}}
+#pragma acc enter data attach(Var)
+
+ // OpenACC TODO: The following 'enter data' directives should diagnose, since
+ // they don't have at least one of the above clauses.
+
+ // expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
+#pragma acc enter data if(Var)
+ // expected-warning at +1{{OpenACC clause 'async' not yet implemented}}
+#pragma acc enter data async
+ // expected-warning at +1{{OpenACC clause 'wait' not yet implemented}}
+#pragma acc enter data wait
+#pragma acc enter data
+
+ // Exit Data
+ // expected-warning at +1{{OpenACC clause 'copyout' not yet implemented}}
+#pragma acc exit data copyout(Var)
+ // expected-warning at +1{{OpenACC clause 'delete' not yet implemented}}
+#pragma acc exit data delete(Var)
+ // expected-warning at +1{{OpenACC clause 'detach' not yet implemented}}
+#pragma acc exit data detach(Var)
+
+ // OpenACC TODO: The following 'exit data' directives should diagnose, since
+ // they don't have at least one of the above clauses.
+
+ // expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
+#pragma acc exit data if(Var)
+ // expected-warning at +1{{OpenACC clause 'async' not yet implemented}}
+#pragma acc exit data async
+ // expected-warning at +1{{OpenACC clause 'wait' not yet implemented}}
+#pragma acc exit data wait
+ // expected-warning at +1{{OpenACC clause 'finalize' not yet implemented}}
+#pragma acc exit data finalize
+#pragma acc exit data
+
+ // Host Data
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(Var)
+ ;
+ // OpenACC TODO: The following 'host_data' directives should diagnose, since
+ // they don't have at least one of the above clauses.
+
+ // expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
+#pragma acc host_data if(Var)
+ ;
+ // expected-warning at +1{{OpenACC clause 'if_present' not yet implemented}}
+#pragma acc host_data if_present
+ ;
+#pragma acc host_data
+ ;
+}
+
+void DataRules() {
+ int Var;
+ // OpenACC TODO: Only 'async' and 'wait' are permitted after a device_type, so
+ // the rest of these should diagnose.
+
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'copy' not yet implemented}}
+#pragma acc data device_type(*) copy(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'copyin' not yet implemented}}
+#pragma acc data device_type(*) copyin(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'copyout' not yet implemented}}
+#pragma acc data device_type(*) copyout(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'create' not yet implemented}}
+#pragma acc data device_type(*) create(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'no_create' not yet implemented}}
+#pragma acc data device_type(*) no_create(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'present' not yet implemented}}
+#pragma acc data device_type(*) present(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'deviceptr' not yet implemented}}
+#pragma acc data device_type(*) deviceptr(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'attach' not yet implemented}}
+#pragma acc data device_type(*) attach(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'default' not yet implemented}}
+#pragma acc data device_type(*) default(none)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
+#pragma acc data device_type(*) if(Var)
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'async' not yet implemented}}
+#pragma acc data device_type(*) async
+ ;
+ // expected-warning at +2{{OpenACC clause 'device_type' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'wait' not yet implemented}}
+#pragma acc data device_type(*) wait
+ ;
+}
+
+struct HasMembers {
+ int Member;
+
+ void HostDataError() {
+ // TODO OpenACC: The following 3 should error, as use_device's var only allows
+ // a variable or array, not an array index, or sub expression.
+
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(this)
+ ;
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(this->Member)
+ ;
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(Member)
+ ;
+ }
+};
+
+void HostDataRules() {
+ int Var, Var2;
+ // TODO OpenACC: The following line should diagnose, since only 1 'if' is
+ // allowed per directive on host_data.
+ // expected-warning at +2{{OpenACC clause 'if' not yet implemented}}
+ // expected-warning at +1{{OpenACC clause 'if' not yet implemented}}
+#pragma acc host_data if(Var) if (Var2)
+ ;
+
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(Var)
+ ;
+
+ int Array[5];
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(Array)
+ ;
+
+ // TODO OpenACC: The following 3 should error, as use_device's var only allows
+ // a variable or array, not an array index, or sub expression.
+
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(Array[1:1])
+ ;
+
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(Array[1])
+ ;
+ HasMembers HM;
+ // expected-warning at +1{{OpenACC clause 'use_device' not yet implemented}}
+#pragma acc host_data use_device(HM.Member)
+ ;
+
+}
diff --git a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
index dc954e36d765da..b401dd891629aa 100644
--- a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
+++ b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
@@ -323,14 +323,15 @@ void no_other_directives() {
#pragma acc loop collapse(2)
for(unsigned i = 0; i < 5; ++i) {
for(unsigned j = 0; j < 5; ++j) {
-#pragma acc data // expected-warning{{OpenACC construct 'data' not yet implemented}}
+#pragma acc data
+ ;
}
}
// expected-note at +1{{active 'collapse' clause defined here}}
#pragma acc loop collapse(2)
for(unsigned i = 0; i < 5; ++i) {
// expected-error at +1{{OpenACC 'data' construct cannot appear in intervening code of a 'loop' with a 'collapse' clause}}
-#pragma acc data // expected-warning{{OpenACC construct 'data' not yet implemented}}
+#pragma acc data
for(unsigned j = 0; j < 5; ++j) {
}
}
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index 221f419861af10..d0fc69af7c847d 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2185,6 +2185,10 @@ class EnqueueVisitor : public ConstStmtVisitor<EnqueueVisitor, void>,
void VisitOpenACCComputeConstruct(const OpenACCComputeConstruct *D);
void VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *D);
void VisitOpenACCCombinedConstruct(const OpenACCCombinedConstruct *D);
+ void VisitOpenACCDataConstruct(const OpenACCDataConstruct *D);
+ void VisitOpenACCEnterDataConstruct(const OpenACCEnterDataConstruct *D);
+ void VisitOpenACCExitDataConstruct(const OpenACCExitDataConstruct *D);
+ void VisitOpenACCHostDataConstruct(const OpenACCHostDataConstruct *D);
void VisitOMPExecutableDirective(const OMPExecutableDirective *D);
void VisitOMPLoopBasedDirective(const OMPLoopBasedDirective *D);
void VisitOMPLoopDirective(const OMPLoopDirective *D);
@@ -3587,6 +3591,29 @@ void EnqueueVisitor::VisitOpenACCCombinedConstruct(
for (auto *Clause : C->clauses())
EnqueueChildren(Clause);
}
+void EnqueueVisitor::VisitOpenACCDataConstruct(const OpenACCDataConstruct *C) {
+ EnqueueChildren(C);
+ for (auto *Clause : C->clauses())
+ EnqueueChildren(Clause);
+}
+void EnqueueVisitor::VisitOpenACCEnterDataConstruct(
+ const OpenACCEnterDataConstruct *C) {
+ EnqueueChildren(C);
+ for (auto *Clause : C->clauses())
+ EnqueueChildren(Clause);
+}
+void EnqueueVisitor::VisitOpenACCExitDataConstruct(
+ const OpenACCExitDataConstruct *C) {
+ EnqueueChildren(C);
+ for (auto *Clause : C->clauses())
+ EnqueueChildren(Clause);
+}
+void EnqueueVisitor::VisitOpenACCHostDataConstruct(
+ const OpenACCHostDataConstruct *C) {
+ EnqueueChildren(C);
+ for (auto *Clause : C->clauses())
+ EnqueueChildren(Clause);
+}
void EnqueueVisitor::VisitAnnotateAttr(const AnnotateAttr *A) {
EnqueueChildren(A);
@@ -6342,6 +6369,14 @@ CXString clang_getCursorKindSpelling(enum CXCursorKind Kind) {
return cxstring::createRef("OpenACCLoopConstruct");
case CXCursor_OpenACCCombinedConstruct:
return cxstring::createRef("OpenACCCombinedConstruct");
+ case CXCursor_OpenACCDataConstruct:
+ return cxstring::createRef("OpenACCDataConstruct");
+ case CXCursor_OpenACCEnterDataConstruct:
+ return cxstring::createRef("OpenACCEnterDataConstruct");
+ case CXCursor_OpenACCExitDataConstruct:
+ return cxstring::createRef("OpenACCExitDataConstruct");
+ case CXCursor_OpenACCHostDataConstruct:
+ return cxstring::createRef("OpenACCHostDataConstruct");
}
llvm_unreachable("Unhandled CXCursorKind");
diff --git a/clang/tools/libclang/CXCursor.cpp b/clang/tools/libclang/CXCursor.cpp
index c8cf51d8061328..26935c45ce5f83 100644
--- a/clang/tools/libclang/CXCursor.cpp
+++ b/clang/tools/libclang/CXCursor.cpp
@@ -888,6 +888,18 @@ CXCursor cxcursor::MakeCXCursor(const Stmt *S, const Decl *Parent,
case Stmt::OpenACCCombinedConstructClass:
K = CXCursor_OpenACCCombinedConstruct;
break;
+ case Stmt::OpenACCDataConstructClass:
+ K = CXCursor_OpenACCDataConstruct;
+ break;
+ case Stmt::OpenACCEnterDataConstructClass:
+ K = CXCursor_OpenACCEnterDataConstruct;
+ break;
+ case Stmt::OpenACCExitDataConstructClass:
+ K = CXCursor_OpenACCExitDataConstruct;
+ break;
+ case Stmt::OpenACCHostDataConstructClass:
+ K = CXCursor_OpenACCHostDataConstruct;
+ break;
case Stmt::OMPTargetParallelGenericLoopDirectiveClass:
K = CXCursor_OMPTargetParallelGenericLoopDirective;
break;
More information about the cfe-commits
mailing list