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@+2{{OpenACC clause 'finalize' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'finalize' not yet implemented, clause ignored}} #pragma acc enter data finalize - // expected-warning@+3{{OpenACC clause 'finalize' not yet implemented, clause ignored}} // expected-warning@+2{{OpenACC clause 'finalize' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'finalize' not yet implemented, clause ignored}} #pragma acc enter data finalize finalize - // expected-warning@+3{{OpenACC clause 'finalize' not yet implemented, clause ignored}} - // expected-error@+2{{invalid OpenACC clause 'invalid'}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}} + // expected-warning@+2{{OpenACC clause 'finalize' not yet implemented, clause ignored}} + // expected-error@+1{{invalid OpenACC clause 'invalid'}} #pragma acc enter data finalize invalid - // expected-warning@+3{{OpenACC clause 'finalize' not yet implemented, clause ignored}} - // expected-error@+2{{invalid OpenACC clause 'invalid'}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}} + // expected-warning@+2{{OpenACC clause 'finalize' not yet implemented, clause ignored}} + // expected-error@+1{{invalid OpenACC clause 'invalid'}} #pragma acc enter data finalize invalid invalid finalize - // expected-warning@+3{{OpenACC clause 'wait' not yet implemented, clause ignored}} - // expected-warning@+2{{OpenACC clause 'finalize' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}} + // expected-warning@+2{{OpenACC clause 'wait' not yet implemented, clause ignored}} + // expected-warning@+1{{OpenACC clause 'finalize' not yet implemented, clause ignored}} #pragma acc enter data wait finalize - // expected-warning@+2{{OpenACC clause 'if_present' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'if_present' not yet implemented, clause ignored}} #pragma acc host_data if_present - // expected-warning@+3{{OpenACC clause 'if_present' not yet implemented, clause ignored}} // expected-warning@+2{{OpenACC clause 'if_present' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'if_present' not yet implemented, clause ignored}} #pragma acc host_data if_present, if_present // expected-error@+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@+4{{OpenACC construct 'exit data' not yet implemented}} // expected-error@+3{{expected ','}} // expected-warning@+2{{OpenACC clause 'delete' not yet implemented, clause ignored}} // expected-warning@+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@+3{{OpenACC construct 'exit data' not yet implemented}} // expected-warning@+2{{OpenACC clause 'delete' not yet implemented, clause ignored}} // expected-warning@+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@+4{{OpenACC construct 'exit data' not yet implemented}} // expected-error@+3{{expected ','}} // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented, clause ignored}} // expected-warning@+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@+3{{OpenACC construct 'exit data' not yet implemented}} // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented, clause ignored}} // expected-warning@+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@+2{{OpenACC construct 'exit data' not yet implemented}} // expected-warning@+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@+1{{invalid OpenACC clause 'clause'}} #pragma acc kernels clause list for(;;){} - // expected-error@+2{{invalid OpenACC clause 'clause'}} - // expected-warning@+1{{OpenACC construct 'data' not yet implemented, pragma ignored}} + // expected-error@+1{{invalid OpenACC clause 'clause'}} #pragma acc data clause list for(;;){} - // expected-error@+2{{invalid OpenACC clause 'clause'}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented, pragma ignored}} + // expected-error@+1{{invalid OpenACC clause 'clause'}} #pragma acc enter data clause list for(;;){} - // expected-error@+2{{invalid OpenACC clause 'clause'}} - // expected-warning@+1{{OpenACC construct 'exit data' not yet implemented, pragma ignored}} + // expected-error@+1{{invalid OpenACC clause 'clause'}} #pragma acc exit data clause list for(;;){} // expected-error@+1{{invalid OpenACC directive 'enter invalid'}} @@ -78,8 +75,7 @@ void func() { // expected-error@+1{{expected identifier}} #pragma acc exit } for(;;){} - // expected-error@+2{{invalid OpenACC clause 'clause'}} - // expected-warning@+1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}} + // expected-error@+1{{invalid OpenACC clause 'clause'}} #pragma acc host_data clause list for(;;){} // expected-error@+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@+1{{active 'collapse' clause defined here}} #pragma acc kernels loop collapse(2) for(unsigned i = 0; i < 5; ++i) { // expected-error@+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@+2{{OpenACC construct 'data' not yet implemented}} // expected-warning@+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@+2{{OpenACC construct 'data' not yet implemented}} // expected-warning@+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@+2{{OpenACC construct 'data' not yet implemented}} // expected-warning@+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@+2{{OpenACC 'device_type' clause is not valid on 'enter data' directive}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented}} + // expected-error@+1{{OpenACC 'device_type' clause is not valid on 'enter data' directive}} #pragma acc enter data device_type(I) - // expected-error@+2{{OpenACC 'dtype' clause is not valid on 'enter data' directive}} - // expected-warning@+1{{OpenACC construct 'enter data' not yet implemented}} + // expected-error@+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@+2{{OpenACC construct 'data' not yet implemented}} // expected-warning@+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@+2{{expected statement}} +#pragma acc data + } + { + // expected-error@+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@+1{{OpenACC clause 'copy' not yet implemented}} +#pragma acc data copy(Var) + ; + // expected-warning@+1{{OpenACC clause 'copyin' not yet implemented}} +#pragma acc data copyin(Var) + ; + // expected-warning@+1{{OpenACC clause 'copyout' not yet implemented}} +#pragma acc data copyout(Var) + ; + // expected-warning@+1{{OpenACC clause 'create' not yet implemented}} +#pragma acc data create(Var) + ; + // expected-warning@+1{{OpenACC clause 'no_create' not yet implemented}} +#pragma acc data no_create(Var) + ; + // expected-warning@+1{{OpenACC clause 'present' not yet implemented}} +#pragma acc data present(Var) + ; + // expected-warning@+1{{OpenACC clause 'deviceptr' not yet implemented}} +#pragma acc data deviceptr(Var) + ; + // expected-warning@+1{{OpenACC clause 'attach' not yet implemented}} +#pragma acc data attach(Var) + ; + // expected-warning@+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@+1{{OpenACC clause 'if' not yet implemented}} +#pragma acc data if(Var) + ; + + // expected-warning@+1{{OpenACC clause 'async' not yet implemented}} +#pragma acc data async + ; + + // expected-warning@+1{{OpenACC clause 'wait' not yet implemented}} +#pragma acc data wait + ; + + // expected-warning@+1{{OpenACC clause 'device_type' not yet implemented}} +#pragma acc data device_type(*) + ; +#pragma acc data + ; + + // Enter Data + // expected-warning@+1{{OpenACC clause 'copyin' not yet implemented}} +#pragma acc enter data copyin(Var) + // expected-warning@+1{{OpenACC clause 'create' not yet implemented}} +#pragma acc enter data create(Var) + // expected-warning@+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@+1{{OpenACC clause 'if' not yet implemented}} +#pragma acc enter data if(Var) + // expected-warning@+1{{OpenACC clause 'async' not yet implemented}} +#pragma acc enter data async + // expected-warning@+1{{OpenACC clause 'wait' not yet implemented}} +#pragma acc enter data wait +#pragma acc enter data + + // Exit Data + // expected-warning@+1{{OpenACC clause 'copyout' not yet implemented}} +#pragma acc exit data copyout(Var) + // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} +#pragma acc exit data delete(Var) + // expected-warning@+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@+1{{OpenACC clause 'if' not yet implemented}} +#pragma acc exit data if(Var) + // expected-warning@+1{{OpenACC clause 'async' not yet implemented}} +#pragma acc exit data async + // expected-warning@+1{{OpenACC clause 'wait' not yet implemented}} +#pragma acc exit data wait + // expected-warning@+1{{OpenACC clause 'finalize' not yet implemented}} +#pragma acc exit data finalize +#pragma acc exit data + + // Host Data + // expected-warning@+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@+1{{OpenACC clause 'if' not yet implemented}} +#pragma acc host_data if(Var) + ; + // expected-warning@+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@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'copy' not yet implemented}} +#pragma acc data device_type(*) copy(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'copyin' not yet implemented}} +#pragma acc data device_type(*) copyin(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'copyout' not yet implemented}} +#pragma acc data device_type(*) copyout(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'create' not yet implemented}} +#pragma acc data device_type(*) create(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'no_create' not yet implemented}} +#pragma acc data device_type(*) no_create(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'present' not yet implemented}} +#pragma acc data device_type(*) present(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'deviceptr' not yet implemented}} +#pragma acc data device_type(*) deviceptr(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'attach' not yet implemented}} +#pragma acc data device_type(*) attach(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'default' not yet implemented}} +#pragma acc data device_type(*) default(none) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'if' not yet implemented}} +#pragma acc data device_type(*) if(Var) + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'async' not yet implemented}} +#pragma acc data device_type(*) async + ; + // expected-warning@+2{{OpenACC clause 'device_type' not yet implemented}} + // expected-warning@+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@+1{{OpenACC clause 'use_device' not yet implemented}} +#pragma acc host_data use_device(this) + ; + // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} +#pragma acc host_data use_device(this->Member) + ; + // expected-warning@+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@+2{{OpenACC clause 'if' not yet implemented}} + // expected-warning@+1{{OpenACC clause 'if' not yet implemented}} +#pragma acc host_data if(Var) if (Var2) + ; + + // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} +#pragma acc host_data use_device(Var) + ; + + int Array[5]; + // expected-warning@+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@+1{{OpenACC clause 'use_device' not yet implemented}} +#pragma acc host_data use_device(Array[1:1]) + ; + + // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} +#pragma acc host_data use_device(Array[1]) + ; + HasMembers HM; + // expected-warning@+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@+1{{active 'collapse' clause defined here}} #pragma acc loop collapse(2) for(unsigned i = 0; i < 5; ++i) { // expected-error@+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; _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
