Author: Erich Keane Date: 2024-06-05T06:21:48-07:00 New Revision: 42f4e505a38480b6a714b503dd946ffff31ae029
URL: https://github.com/llvm/llvm-project/commit/42f4e505a38480b6a714b503dd946ffff31ae029 DIFF: https://github.com/llvm/llvm-project/commit/42f4e505a38480b6a714b503dd946ffff31ae029.diff LOG: [OpenACC] Loop construct basic Sema and AST work (#93742) This patch implements the 'loop' construct AST, as well as the basic appertainment rule. Additionally, it sets up the 'parent' compute construct, which is necessary for codegen/other diagnostics. A 'loop' can apply to a for or range-for loop, otherwise it has no other restrictions (though some of its clauses do). Added: clang/test/AST/ast-print-openacc-loop-construct.cpp clang/test/SemaOpenACC/loop-ast.cpp clang/test/SemaOpenACC/loop-loc-and-stmt.c clang/test/SemaOpenACC/loop-loc-and-stmt.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/DiagnosticSemaKinds.td clang/include/clang/Basic/StmtNodes.td clang/include/clang/Sema/SemaOpenACC.h 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/compute-construct-async-clause.c clang/test/SemaOpenACC/compute-construct-attach-clause.c clang/test/SemaOpenACC/compute-construct-copy-clause.c clang/test/SemaOpenACC/compute-construct-copyin-clause.c clang/test/SemaOpenACC/compute-construct-copyout-clause.c clang/test/SemaOpenACC/compute-construct-create-clause.c clang/test/SemaOpenACC/compute-construct-default-clause.c clang/test/SemaOpenACC/compute-construct-deviceptr-clause.c clang/test/SemaOpenACC/compute-construct-firstprivate-clause.c clang/test/SemaOpenACC/compute-construct-if-clause.c clang/test/SemaOpenACC/compute-construct-no_create-clause.c clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c clang/test/SemaOpenACC/compute-construct-num_workers-clause.c clang/test/SemaOpenACC/compute-construct-present-clause.c clang/test/SemaOpenACC/compute-construct-self-clause.c clang/test/SemaOpenACC/compute-construct-vector_length-clause.c clang/test/SemaOpenACC/compute-construct-wait-clause.c 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 365b607c74117..ce2282937f86c 100644 --- a/clang/include/clang-c/Index.h +++ b/clang/include/clang-c/Index.h @@ -2150,7 +2150,11 @@ enum CXCursorKind { */ CXCursor_OpenACCComputeConstruct = 320, - CXCursor_LastStmt = CXCursor_OpenACCComputeConstruct, + /** OpenACC Loop Construct. + */ + CXCursor_OpenACCLoopConstruct = 321, + + CXCursor_LastStmt = CXCursor_OpenACCLoopConstruct, /** * Cursor that represents the translation unit itself. diff --git a/clang/include/clang/AST/RecursiveASTVisitor.h b/clang/include/clang/AST/RecursiveASTVisitor.h index 99093aa17972c..aa55e2e7e8718 100644 --- a/clang/include/clang/AST/RecursiveASTVisitor.h +++ b/clang/include/clang/AST/RecursiveASTVisitor.h @@ -4000,6 +4000,8 @@ bool RecursiveASTVisitor<Derived>::VisitOpenACCClauseList( DEF_TRAVERSE_STMT(OpenACCComputeConstruct, { TRY_TO(TraverseOpenACCAssociatedStmtConstruct(S)); }) +DEF_TRAVERSE_STMT(OpenACCLoopConstruct, + { TRY_TO(TraverseOpenACCAssociatedStmtConstruct(S)); }) // FIXME: look at the following tricky-seeming exprs to see if we // need to recurse on anything. These are ones that have methods diff --git a/clang/include/clang/AST/StmtOpenACC.h b/clang/include/clang/AST/StmtOpenACC.h index 04daf511f5871..b3aea09be03dd 100644 --- a/clang/include/clang/AST/StmtOpenACC.h +++ b/clang/include/clang/AST/StmtOpenACC.h @@ -113,6 +113,8 @@ class OpenACCAssociatedStmtConstruct : public OpenACCConstructStmt { return const_cast<OpenACCAssociatedStmtConstruct *>(this)->children(); } }; + +class OpenACCLoopConstruct; /// This class represents a compute construct, representing a 'Kind' of /// `parallel', 'serial', or 'kernel'. These constructs are associated with a /// 'structured block', defined as: @@ -165,6 +167,11 @@ class OpenACCComputeConstruct final } void setStructuredBlock(Stmt *S) { setAssociatedStmt(S); } + // Serialization helper function that searches the structured block for 'loop' + // constructs that should be associated with this, and sets their parent + // compute construct to this one. This isn't necessary normally, since we have + // the ability to record the state during parsing. + void findAndSetChildLoops(); public: static bool classof(const Stmt *T) { @@ -176,12 +183,74 @@ class OpenACCComputeConstruct final static OpenACCComputeConstruct * Create(const ASTContext &C, OpenACCDirectiveKind K, SourceLocation BeginLoc, SourceLocation DirectiveLoc, SourceLocation EndLoc, - ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock); + ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock, + ArrayRef<OpenACCLoopConstruct *> AssociatedLoopConstructs); Stmt *getStructuredBlock() { return getAssociatedStmt(); } const Stmt *getStructuredBlock() const { return const_cast<OpenACCComputeConstruct *>(this)->getStructuredBlock(); } }; +/// This class represents a 'loop' construct. The 'loop' construct applies to a +/// 'for' loop (or range-for loop), and is optionally associated with a Compute +/// Construct. +class OpenACCLoopConstruct final + : public OpenACCAssociatedStmtConstruct, + public llvm::TrailingObjects<OpenACCLoopConstruct, + const OpenACCClause *> { + // The compute construct this loop is associated with, or nullptr if this is + // an orphaned loop construct, or if it hasn't been set yet. Because we + // construct the directives at the end of their statement, the 'parent' + // construct is not yet available at the time of construction, so this needs + // to be set 'later'. + const OpenACCComputeConstruct *ParentComputeConstruct = nullptr; + + friend class ASTStmtWriter; + friend class ASTStmtReader; + friend class ASTContext; + friend class OpenACCComputeConstruct; + + OpenACCLoopConstruct(unsigned NumClauses); + + OpenACCLoopConstruct(SourceLocation Start, SourceLocation DirLoc, + SourceLocation End, + ArrayRef<const OpenACCClause *> Clauses, Stmt *Loop); + void setLoop(Stmt *Loop); + + void setParentComputeConstruct(OpenACCComputeConstruct *CC) { + assert(!ParentComputeConstruct && "Parent already set?"); + ParentComputeConstruct = CC; + } + +public: + static bool classof(const Stmt *T) { + return T->getStmtClass() == OpenACCLoopConstructClass; + } + + static OpenACCLoopConstruct *CreateEmpty(const ASTContext &C, + unsigned NumClauses); + + static OpenACCLoopConstruct * + Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation DirLoc, + SourceLocation EndLoc, ArrayRef<const OpenACCClause *> Clauses, + Stmt *Loop); + + Stmt *getLoop() { return getAssociatedStmt(); } + const Stmt *getLoop() const { + return const_cast<OpenACCLoopConstruct *>(this)->getLoop(); + } + + /// OpenACC 3.3 2.9: + /// An orphaned loop construct is a loop construct that is not lexically + /// enclosed within a compute construct. The parent compute construct of a + /// loop construct is the nearest compute construct that lexically contains + /// the loop construct. + bool isOrphanedLoopConstruct() const { + return ParentComputeConstruct == nullptr; + } + const OpenACCComputeConstruct *getParentComputeConstruct() const { + return ParentComputeConstruct; + } +}; } // 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 caa33abd99e47..abfafcaef271b 100644 --- a/clang/include/clang/AST/TextNodeDumper.h +++ b/clang/include/clang/AST/TextNodeDumper.h @@ -408,6 +408,7 @@ class TextNodeDumper VisitLifetimeExtendedTemporaryDecl(const LifetimeExtendedTemporaryDecl *D); void VisitHLSLBufferDecl(const HLSLBufferDecl *D); void VisitOpenACCConstructStmt(const OpenACCConstructStmt *S); + void VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *S); }; } // namespace clang diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td index e34eb692941b4..50332966a7e37 100644 --- a/clang/include/clang/Basic/DiagnosticSemaKinds.td +++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td @@ -12413,6 +12413,9 @@ def err_acc_reduction_composite_type def err_acc_reduction_composite_member_type :Error< "OpenACC 'reduction' composite variable must not have non-scalar field">; def note_acc_reduction_composite_member_loc : Note<"invalid field is here">; +def err_acc_loop_not_for_loop + : Error<"OpenACC 'loop' construct can only be applied to a 'for' loop">; +def note_acc_construct_here : Note<"'%0' construct is here">; // AMDGCN builtins diagnostics def err_amdgcn_global_load_lds_size_invalid_value : Error<"invalid size value">; diff --git a/clang/include/clang/Basic/StmtNodes.td b/clang/include/clang/Basic/StmtNodes.td index 305f19daa4a92..6ca08abdb14f0 100644 --- a/clang/include/clang/Basic/StmtNodes.td +++ b/clang/include/clang/Basic/StmtNodes.td @@ -302,3 +302,4 @@ def OpenACCConstructStmt : StmtNode<Stmt, /*abstract=*/1>; def OpenACCAssociatedStmtConstruct : StmtNode<OpenACCConstructStmt, /*abstract=*/1>; def OpenACCComputeConstruct : StmtNode<OpenACCAssociatedStmtConstruct>; +def OpenACCLoopConstruct : StmtNode<OpenACCAssociatedStmtConstruct>; diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h index 66144de4340a8..a5f2a8bf74657 100644 --- a/clang/include/clang/Sema/SemaOpenACC.h +++ b/clang/include/clang/Sema/SemaOpenACC.h @@ -15,6 +15,7 @@ #define LLVM_CLANG_SEMA_SEMAOPENACC_H #include "clang/AST/DeclGroup.h" +#include "clang/AST/StmtOpenACC.h" #include "clang/Basic/OpenACCKinds.h" #include "clang/Basic/SourceLocation.h" #include "clang/Sema/Ownership.h" @@ -25,6 +26,15 @@ namespace clang { class OpenACCClause; class SemaOpenACC : public SemaBase { +private: + /// A collection of loop constructs in the compute construct scope that + /// haven't had their 'parent' compute construct set yet. Entires will only be + /// made to this list in the case where we know the loop isn't an orphan. + llvm::SmallVector<OpenACCLoopConstruct *> ParentlessLoopConstructs; + /// Whether we are inside of a compute construct, and should add loops to the + /// above collection. + bool InsideComputeConstruct = false; + public: // Redeclaration of the version in OpenACCClause.h. using DeviceTypeArgument = std::pair<IdentifierInfo *, SourceLocation>; @@ -394,7 +404,8 @@ class SemaOpenACC : public SemaBase { bool ActOnStartDeclDirective(OpenACCDirectiveKind K, SourceLocation StartLoc); /// Called when we encounter an associated statement for our construct, this /// should check legality of the statement as it appertains to this Construct. - StmtResult ActOnAssociatedStmt(OpenACCDirectiveKind K, StmtResult AssocStmt); + StmtResult ActOnAssociatedStmt(SourceLocation DirectiveLoc, + OpenACCDirectiveKind K, StmtResult AssocStmt); /// Called after the directive has been completely parsed, including the /// declaration group or associated statement. @@ -431,6 +442,20 @@ class SemaOpenACC : public SemaBase { Expr *LowerBound, SourceLocation ColonLocFirst, Expr *Length, SourceLocation RBLoc); + + /// Helper type for the registration/assignment of constructs that need to + /// 'know' about their parent constructs and hold a reference to them, such as + /// Loop needing its parent construct. + class AssociatedStmtRAII { + SemaOpenACC &SemaRef; + bool WasInsideComputeConstruct; + OpenACCDirectiveKind DirKind; + llvm::SmallVector<OpenACCLoopConstruct *> ParentlessLoopConstructs; + + public: + AssociatedStmtRAII(SemaOpenACC &, OpenACCDirectiveKind); + ~AssociatedStmtRAII(); + }; }; } // namespace clang diff --git a/clang/include/clang/Serialization/ASTBitCodes.h b/clang/include/clang/Serialization/ASTBitCodes.h index fe1bd47348be1..f59ff6af4c764 100644 --- a/clang/include/clang/Serialization/ASTBitCodes.h +++ b/clang/include/clang/Serialization/ASTBitCodes.h @@ -1946,6 +1946,7 @@ enum StmtCode { // OpenACC Constructs STMT_OPENACC_COMPUTE_CONSTRUCT, + STMT_OPENACC_LOOP_CONSTRUCT, }; /// The kinds of designators that can occur in a diff --git a/clang/lib/AST/StmtOpenACC.cpp b/clang/lib/AST/StmtOpenACC.cpp index 47899b344c97a..2d864a2885796 100644 --- a/clang/lib/AST/StmtOpenACC.cpp +++ b/clang/lib/AST/StmtOpenACC.cpp @@ -12,6 +12,8 @@ #include "clang/AST/StmtOpenACC.h" #include "clang/AST/ASTContext.h" +#include "clang/AST/RecursiveASTVisitor.h" +#include "clang/AST/StmtCXX.h" using namespace clang; OpenACCComputeConstruct * @@ -26,11 +28,98 @@ OpenACCComputeConstruct::CreateEmpty(const ASTContext &C, unsigned NumClauses) { OpenACCComputeConstruct *OpenACCComputeConstruct::Create( const ASTContext &C, OpenACCDirectiveKind K, SourceLocation BeginLoc, SourceLocation DirLoc, SourceLocation EndLoc, - ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock) { + ArrayRef<const OpenACCClause *> Clauses, Stmt *StructuredBlock, + ArrayRef<OpenACCLoopConstruct *> AssociatedLoopConstructs) { void *Mem = C.Allocate( OpenACCComputeConstruct::totalSizeToAlloc<const OpenACCClause *>( Clauses.size())); auto *Inst = new (Mem) OpenACCComputeConstruct(K, BeginLoc, DirLoc, EndLoc, Clauses, StructuredBlock); + + llvm::for_each(AssociatedLoopConstructs, [&](OpenACCLoopConstruct *C) { + C->setParentComputeConstruct(Inst); + }); + + return Inst; +} + +void OpenACCComputeConstruct::findAndSetChildLoops() { + struct LoopConstructFinder : RecursiveASTVisitor<LoopConstructFinder> { + OpenACCComputeConstruct *Construct = nullptr; + + LoopConstructFinder(OpenACCComputeConstruct *Construct) + : Construct(Construct) {} + + bool TraverseOpenACCComputeConstruct(OpenACCComputeConstruct *C) { + // Stop searching if we find a compute construct. + return true; + } + bool TraverseOpenACCLoopConstruct(OpenACCLoopConstruct *C) { + // Stop searching if we find a loop construct, after taking ownership of + // it. + C->setParentComputeConstruct(Construct); + return true; + } + }; + + LoopConstructFinder f(this); + f.TraverseStmt(getAssociatedStmt()); +} + +OpenACCLoopConstruct::OpenACCLoopConstruct(unsigned NumClauses) + : OpenACCAssociatedStmtConstruct( + OpenACCLoopConstructClass, OpenACCDirectiveKind::Loop, + SourceLocation{}, SourceLocation{}, SourceLocation{}, + /*AssociatedStmt=*/nullptr) { + std::uninitialized_value_construct( + getTrailingObjects<const OpenACCClause *>(), + getTrailingObjects<const OpenACCClause *>() + NumClauses); + setClauseList( + MutableArrayRef(getTrailingObjects<const OpenACCClause *>(), NumClauses)); +} + +OpenACCLoopConstruct::OpenACCLoopConstruct( + SourceLocation Start, SourceLocation DirLoc, SourceLocation End, + ArrayRef<const OpenACCClause *> Clauses, Stmt *Loop) + : OpenACCAssociatedStmtConstruct(OpenACCLoopConstructClass, + OpenACCDirectiveKind::Loop, Start, DirLoc, + End, Loop) { + // accept 'nullptr' for the loop. This is diagnosed somewhere, but this gives + // us some level of AST fidelity in the error case. + assert((Loop == nullptr || isa<ForStmt, CXXForRangeStmt>(Loop)) && + "Associated Loop not a for loop?"); + // Initialize the trailing storage. + std::uninitialized_copy(Clauses.begin(), Clauses.end(), + getTrailingObjects<const OpenACCClause *>()); + + setClauseList(MutableArrayRef(getTrailingObjects<const OpenACCClause *>(), + Clauses.size())); +} + +void OpenACCLoopConstruct::setLoop(Stmt *Loop) { + assert((isa<ForStmt, CXXForRangeStmt>(Loop)) && + "Associated Loop not a for loop?"); + setAssociatedStmt(Loop); +} + +OpenACCLoopConstruct *OpenACCLoopConstruct::CreateEmpty(const ASTContext &C, + unsigned NumClauses) { + void *Mem = + C.Allocate(OpenACCLoopConstruct::totalSizeToAlloc<const OpenACCClause *>( + NumClauses)); + auto *Inst = new (Mem) OpenACCLoopConstruct(NumClauses); + return Inst; +} + +OpenACCLoopConstruct * +OpenACCLoopConstruct::Create(const ASTContext &C, SourceLocation BeginLoc, + SourceLocation DirLoc, SourceLocation EndLoc, + ArrayRef<const OpenACCClause *> Clauses, + Stmt *Loop) { + void *Mem = + C.Allocate(OpenACCLoopConstruct::totalSizeToAlloc<const OpenACCClause *>( + Clauses.size())); + auto *Inst = + new (Mem) OpenACCLoopConstruct(BeginLoc, DirLoc, EndLoc, Clauses, Loop); return Inst; } diff --git a/clang/lib/AST/StmtPrinter.cpp b/clang/lib/AST/StmtPrinter.cpp index be2d5a2eb6b46..7e030e0551269 100644 --- a/clang/lib/AST/StmtPrinter.cpp +++ b/clang/lib/AST/StmtPrinter.cpp @@ -1156,6 +1156,19 @@ void StmtPrinter::VisitOpenACCComputeConstruct(OpenACCComputeConstruct *S) { PrintStmt(S->getStructuredBlock()); } +void StmtPrinter::VisitOpenACCLoopConstruct(OpenACCLoopConstruct *S) { + Indent() << "#pragma acc loop"; + + if (!S->clauses().empty()) { + OS << ' '; + OpenACCClausePrinter Printer(OS, Policy); + Printer.VisitClauseList(S->clauses()); + } + OS << '\n'; + + PrintStmt(S->getLoop()); +} + //===----------------------------------------------------------------------===// // Expr printing methods. //===----------------------------------------------------------------------===// diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp index 00b8c43af035c..6d9a76120cfef 100644 --- a/clang/lib/AST/StmtProfile.cpp +++ b/clang/lib/AST/StmtProfile.cpp @@ -2605,6 +2605,14 @@ void StmtProfiler::VisitOpenACCComputeConstruct( P.VisitOpenACCClauseList(S->clauses()); } +void StmtProfiler::VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *S) { + // VisitStmt handles children, so the Loop is handled. + VisitStmt(S); + + OpenACCClauseProfiler P{*this}; + P.VisitOpenACCClauseList(S->clauses()); +} + void Stmt::Profile(llvm::FoldingSetNodeID &ID, const ASTContext &Context, bool Canonical, bool ProfileLambdaExpr) const { StmtProfilerWithPointers Profiler(ID, Context, Canonical, ProfileLambdaExpr); diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp index 0e0e0a86f5cfc..b2bf259ec24ee 100644 --- a/clang/lib/AST/TextNodeDumper.cpp +++ b/clang/lib/AST/TextNodeDumper.cpp @@ -2869,3 +2869,10 @@ void TextNodeDumper::VisitHLSLBufferDecl(const HLSLBufferDecl *D) { void TextNodeDumper::VisitOpenACCConstructStmt(const OpenACCConstructStmt *S) { OS << " " << S->getDirectiveKind(); } +void TextNodeDumper::VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *S) { + + if (S->isOrphanedLoopConstruct()) + OS << " <orphan>"; + else + OS << " parent: " << S->getParentComputeConstruct(); +} diff --git a/clang/lib/CodeGen/CGStmt.cpp b/clang/lib/CodeGen/CGStmt.cpp index 99daaa14cf3fe..41ac511c52f51 100644 --- a/clang/lib/CodeGen/CGStmt.cpp +++ b/clang/lib/CodeGen/CGStmt.cpp @@ -442,6 +442,9 @@ void CodeGenFunction::EmitStmt(const Stmt *S, ArrayRef<const Attr *> Attrs) { case Stmt::OpenACCComputeConstructClass: EmitOpenACCComputeConstruct(cast<OpenACCComputeConstruct>(*S)); break; + case Stmt::OpenACCLoopConstructClass: + EmitOpenACCLoopConstruct(cast<OpenACCLoopConstruct>(*S)); + break; } } diff --git a/clang/lib/CodeGen/CodeGenFunction.h b/clang/lib/CodeGen/CodeGenFunction.h index 45585361a4fc9..5739fbaaa9194 100644 --- a/clang/lib/CodeGen/CodeGenFunction.h +++ b/clang/lib/CodeGen/CodeGenFunction.h @@ -4062,6 +4062,13 @@ class CodeGenFunction : public CodeGenTypeCache { EmitStmt(S.getStructuredBlock()); } + void EmitOpenACCLoopConstruct(const OpenACCLoopConstruct &S) { + // TODO OpenACC: Implement this. It is currently implemented as a 'no-op', + // simply emitting its loop, but in the future we will implement + // some sort of IR. + EmitStmt(S.getLoop()); + } + //===--------------------------------------------------------------------===// // LValue Expression Emission //===--------------------------------------------------------------------===// diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp index 63afc18783a1f..c7b6763b4dbdd 100644 --- a/clang/lib/Parse/ParseOpenACC.cpp +++ b/clang/lib/Parse/ParseOpenACC.cpp @@ -571,6 +571,7 @@ bool doesDirectiveHaveAssociatedStmt(OpenACCDirectiveKind DirKind) { case OpenACCDirectiveKind::Parallel: case OpenACCDirectiveKind::Serial: case OpenACCDirectiveKind::Kernels: + case OpenACCDirectiveKind::Loop: return true; } llvm_unreachable("Unhandled directive->assoc stmt"); @@ -1447,13 +1448,14 @@ StmtResult Parser::ParseOpenACCDirectiveStmt() { return StmtError(); StmtResult AssocStmt; - + SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(getActions().OpenACC(), + DirInfo.DirKind); if (doesDirectiveHaveAssociatedStmt(DirInfo.DirKind)) { ParsingOpenACCDirectiveRAII DirScope(*this, /*Value=*/false); ParseScope ACCScope(this, getOpenACCScopeFlags(DirInfo.DirKind)); - AssocStmt = getActions().OpenACC().ActOnAssociatedStmt(DirInfo.DirKind, - ParseStatement()); + AssocStmt = getActions().OpenACC().ActOnAssociatedStmt( + DirInfo.StartLoc, DirInfo.DirKind, ParseStatement()); } return getActions().OpenACC().ActOnEndStmtDirective( diff --git a/clang/lib/Sema/SemaExceptionSpec.cpp b/clang/lib/Sema/SemaExceptionSpec.cpp index 41bf273d12f2f..17acfca6b0112 100644 --- a/clang/lib/Sema/SemaExceptionSpec.cpp +++ b/clang/lib/Sema/SemaExceptionSpec.cpp @@ -1425,6 +1425,7 @@ CanThrowResult Sema::canThrow(const Stmt *S) { // Most statements can throw if any substatement can throw. case Stmt::OpenACCComputeConstructClass: + case Stmt::OpenACCLoopConstructClass: case Stmt::AttributedStmtClass: case Stmt::BreakStmtClass: case Stmt::CapturedStmtClass: diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp index 5b4d01860c0bd..92da218010c94 100644 --- a/clang/lib/Sema/SemaOpenACC.cpp +++ b/clang/lib/Sema/SemaOpenACC.cpp @@ -33,6 +33,7 @@ bool diagnoseConstructAppertainment(SemaOpenACC &S, OpenACCDirectiveKind K, case OpenACCDirectiveKind::Parallel: case OpenACCDirectiveKind::Serial: case OpenACCDirectiveKind::Kernels: + case OpenACCDirectiveKind::Loop: if (!IsStmt) return S.Diag(StartLoc, diag::err_acc_construct_appertainment) << K; break; @@ -370,6 +371,30 @@ bool checkValidAfterDeviceType( SemaOpenACC::SemaOpenACC(Sema &S) : SemaBase(S) {} +SemaOpenACC::AssociatedStmtRAII::AssociatedStmtRAII(SemaOpenACC &S, + OpenACCDirectiveKind DK) + : SemaRef(S), WasInsideComputeConstruct(S.InsideComputeConstruct), + DirKind(DK) { + // Compute constructs end up taking their 'loop'. + if (DirKind == OpenACCDirectiveKind::Parallel || + DirKind == OpenACCDirectiveKind::Serial || + DirKind == OpenACCDirectiveKind::Kernels) { + SemaRef.InsideComputeConstruct = true; + SemaRef.ParentlessLoopConstructs.swap(ParentlessLoopConstructs); + } +} + +SemaOpenACC::AssociatedStmtRAII::~AssociatedStmtRAII() { + SemaRef.InsideComputeConstruct = WasInsideComputeConstruct; + if (DirKind == OpenACCDirectiveKind::Parallel || + DirKind == OpenACCDirectiveKind::Serial || + DirKind == OpenACCDirectiveKind::Kernels) { + assert(SemaRef.ParentlessLoopConstructs.empty() && + "Didn't consume loop construct list?"); + SemaRef.ParentlessLoopConstructs.swap(ParentlessLoopConstructs); + } +} + OpenACCClause * SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses, OpenACCParsedClause &Clause) { @@ -927,6 +952,7 @@ void SemaOpenACC::ActOnConstruct(OpenACCDirectiveKind K, case OpenACCDirectiveKind::Parallel: case OpenACCDirectiveKind::Serial: case OpenACCDirectiveKind::Kernels: + case OpenACCDirectiveKind::Loop: // Nothing to do here, there is no real legalization that needs to happen // here as these constructs do not take any arguments. break; @@ -1348,16 +1374,34 @@ StmtResult SemaOpenACC::ActOnEndStmtDirective(OpenACCDirectiveKind K, return StmtError(); case OpenACCDirectiveKind::Parallel: case OpenACCDirectiveKind::Serial: - case OpenACCDirectiveKind::Kernels: - // TODO OpenACC: Add clauses to the construct here. - return OpenACCComputeConstruct::Create( + case OpenACCDirectiveKind::Kernels: { + auto *ComputeConstruct = OpenACCComputeConstruct::Create( getASTContext(), K, StartLoc, DirLoc, EndLoc, Clauses, + AssocStmt.isUsable() ? AssocStmt.get() : nullptr, + ParentlessLoopConstructs); + + ParentlessLoopConstructs.clear(); + return ComputeConstruct; + } + case OpenACCDirectiveKind::Loop: { + auto *LoopConstruct = OpenACCLoopConstruct::Create( + getASTContext(), StartLoc, DirLoc, EndLoc, Clauses, AssocStmt.isUsable() ? AssocStmt.get() : nullptr); + + // If we are in the scope of a compute construct, add this to the list of + // loop constructs that need assigning to the next closing compute + // construct. + if (InsideComputeConstruct) + ParentlessLoopConstructs.push_back(LoopConstruct); + + return LoopConstruct; + } } llvm_unreachable("Unhandled case in directive handling?"); } -StmtResult SemaOpenACC::ActOnAssociatedStmt(OpenACCDirectiveKind K, +StmtResult SemaOpenACC::ActOnAssociatedStmt(SourceLocation DirectiveLoc, + OpenACCDirectiveKind K, StmtResult AssocStmt) { switch (K) { default: @@ -1375,6 +1419,14 @@ StmtResult SemaOpenACC::ActOnAssociatedStmt(OpenACCDirectiveKind K, // an interpretation of it is to allow this and treat the initializer as // the 'structured block'. return AssocStmt; + case OpenACCDirectiveKind::Loop: + if (AssocStmt.isUsable() && + !isa<CXXForRangeStmt, ForStmt>(AssocStmt.get())) { + Diag(AssocStmt.get()->getBeginLoc(), diag::err_acc_loop_not_for_loop); + Diag(DirectiveLoc, diag::note_acc_construct_here) << K; + return StmtError(); + } + return AssocStmt; } llvm_unreachable("Invalid associated statement application"); } diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h index 70603ba6c2717..07f995edaf3d4 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -4041,6 +4041,15 @@ class TreeTransform { EndLoc, Clauses, StrBlock); } + StmtResult RebuildOpenACCLoopConstruct(SourceLocation BeginLoc, + SourceLocation DirLoc, + SourceLocation EndLoc, + ArrayRef<OpenACCClause *> Clauses, + StmtResult Loop) { + return getSema().OpenACC().ActOnEndStmtDirective( + OpenACCDirectiveKind::Loop, BeginLoc, DirLoc, EndLoc, Clauses, Loop); + } + private: TypeLoc TransformTypeInObjectScope(TypeLoc TL, QualType ObjectType, @@ -11541,8 +11550,6 @@ template <typename Derived> StmtResult TreeTransform<Derived>::TransformOpenACCComputeConstruct( OpenACCComputeConstruct *C) { getSema().OpenACC().ActOnConstruct(C->getDirectiveKind(), C->getBeginLoc()); - // FIXME: When implementing this for constructs that can take arguments, we - // should do Sema for them here. if (getSema().OpenACC().ActOnStartStmtDirective(C->getDirectiveKind(), C->getBeginLoc())) @@ -11551,17 +11558,44 @@ StmtResult TreeTransform<Derived>::TransformOpenACCComputeConstruct( llvm::SmallVector<OpenACCClause *> TransformedClauses = getDerived().TransformOpenACCClauseList(C->getDirectiveKind(), C->clauses()); - // Transform Structured Block. + SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(getSema().OpenACC(), + C->getDirectiveKind()); StmtResult StrBlock = getDerived().TransformStmt(C->getStructuredBlock()); - StrBlock = - getSema().OpenACC().ActOnAssociatedStmt(C->getDirectiveKind(), StrBlock); + StrBlock = getSema().OpenACC().ActOnAssociatedStmt( + C->getBeginLoc(), C->getDirectiveKind(), StrBlock); return getDerived().RebuildOpenACCComputeConstruct( C->getDirectiveKind(), C->getBeginLoc(), C->getDirectiveLoc(), C->getEndLoc(), TransformedClauses, StrBlock); } +template <typename Derived> +StmtResult +TreeTransform<Derived>::TransformOpenACCLoopConstruct(OpenACCLoopConstruct *C) { + + getSema().OpenACC().ActOnConstruct(C->getDirectiveKind(), C->getBeginLoc()); + + if (getSema().OpenACC().ActOnStartStmtDirective(C->getDirectiveKind(), + C->getBeginLoc())) + return StmtError(); + + llvm::SmallVector<OpenACCClause *> TransformedClauses = + getDerived().TransformOpenACCClauseList(C->getDirectiveKind(), + C->clauses()); + + // Transform Loop. + SemaOpenACC::AssociatedStmtRAII AssocStmtRAII(getSema().OpenACC(), + C->getDirectiveKind()); + StmtResult Loop = getDerived().TransformStmt(C->getLoop()); + Loop = getSema().OpenACC().ActOnAssociatedStmt(C->getBeginLoc(), + C->getDirectiveKind(), Loop); + + return getDerived().RebuildOpenACCLoopConstruct( + C->getBeginLoc(), C->getDirectiveLoc(), C->getEndLoc(), + TransformedClauses, Loop); +} + //===----------------------------------------------------------------------===// // Expression transformation //===----------------------------------------------------------------------===// diff --git a/clang/lib/Serialization/ASTReaderStmt.cpp b/clang/lib/Serialization/ASTReaderStmt.cpp index bea2b94989107..67ef170251914 100644 --- a/clang/lib/Serialization/ASTReaderStmt.cpp +++ b/clang/lib/Serialization/ASTReaderStmt.cpp @@ -2810,6 +2810,12 @@ void ASTStmtReader::VisitOpenACCAssociatedStmtConstruct( void ASTStmtReader::VisitOpenACCComputeConstruct(OpenACCComputeConstruct *S) { VisitStmt(S); VisitOpenACCAssociatedStmtConstruct(S); + S->findAndSetChildLoops(); +} + +void ASTStmtReader::VisitOpenACCLoopConstruct(OpenACCLoopConstruct *S) { + VisitStmt(S); + VisitOpenACCAssociatedStmtConstruct(S); } //===----------------------------------------------------------------------===// @@ -4235,6 +4241,11 @@ Stmt *ASTReader::ReadStmtFromStream(ModuleFile &F) { S = OpenACCComputeConstruct::CreateEmpty(Context, NumClauses); break; } + case STMT_OPENACC_LOOP_CONSTRUCT: { + unsigned NumClauses = Record[ASTStmtReader::NumStmtFields]; + S = OpenACCLoopConstruct::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 3c586b270fbf4..1a98e30e0f89f 100644 --- a/clang/lib/Serialization/ASTWriterStmt.cpp +++ b/clang/lib/Serialization/ASTWriterStmt.cpp @@ -2863,6 +2863,12 @@ void ASTStmtWriter::VisitOpenACCComputeConstruct(OpenACCComputeConstruct *S) { Code = serialization::STMT_OPENACC_COMPUTE_CONSTRUCT; } +void ASTStmtWriter::VisitOpenACCLoopConstruct(OpenACCLoopConstruct *S) { + VisitStmt(S); + VisitOpenACCAssociatedStmtConstruct(S); + Code = serialization::STMT_OPENACC_LOOP_CONSTRUCT; +} + //===----------------------------------------------------------------------===// // ASTWriter Implementation //===----------------------------------------------------------------------===// diff --git a/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp b/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp index 793f3a63ea29e..290d96611d466 100644 --- a/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp +++ b/clang/lib/StaticAnalyzer/Core/ExprEngine.cpp @@ -1822,6 +1822,7 @@ void ExprEngine::Visit(const Stmt *S, ExplodedNode *Pred, case Stmt::OMPTargetParallelGenericLoopDirectiveClass: case Stmt::CapturedStmtClass: case Stmt::OpenACCComputeConstructClass: + case Stmt::OpenACCLoopConstructClass: case Stmt::OMPUnrollDirectiveClass: case Stmt::OMPMetaDirectiveClass: { const ExplodedNode *node = Bldr.generateSink(S, Pred, Pred->getState()); diff --git a/clang/test/AST/ast-print-openacc-loop-construct.cpp b/clang/test/AST/ast-print-openacc-loop-construct.cpp new file mode 100644 index 0000000000000..21c92b17317ef --- /dev/null +++ b/clang/test/AST/ast-print-openacc-loop-construct.cpp @@ -0,0 +1,9 @@ +// RUN: %clang_cc1 -fopenacc -Wno-openacc-deprecated-clause-alias -ast-print %s -o - | FileCheck %s + +void foo() { +// CHECK: #pragma acc loop +// CHECK-NEXT: for (;;) +// CHECK-NEXT: ; +#pragma acc loop + for(;;); +} diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c index 49e749feb2ec7..cb118f69fb447 100644 --- a/clang/test/ParserOpenACC/parse-clauses.c +++ b/clang/test/ParserOpenACC/parse-clauses.c @@ -37,23 +37,23 @@ void func() { // expected-warning@+1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}} #pragma acc host_data if_present, if_present - // expected-warning@+4{{OpenACC clause 'seq' not yet implemented, clause ignored}} - // expected-warning@+3{{OpenACC clause 'independent' not yet implemented, clause ignored}} - // expected-warning@+2{{OpenACC clause 'auto' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+3{{OpenACC clause 'seq' not yet implemented, clause ignored}} + // expected-warning@+2{{OpenACC clause 'independent' not yet implemented, clause ignored}} + // expected-warning@+1{{OpenACC clause 'auto' not yet implemented, clause ignored}} #pragma acc loop seq independent auto + for(;;){} - // expected-warning@+4{{OpenACC clause 'seq' not yet implemented, clause ignored}} - // expected-warning@+3{{OpenACC clause 'independent' not yet implemented, clause ignored}} - // expected-warning@+2{{OpenACC clause 'auto' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+3{{OpenACC clause 'seq' not yet implemented, clause ignored}} + // expected-warning@+2{{OpenACC clause 'independent' not yet implemented, clause ignored}} + // expected-warning@+1{{OpenACC clause 'auto' not yet implemented, clause ignored}} #pragma acc loop seq, independent auto + for(;;){} - // expected-warning@+4{{OpenACC clause 'seq' not yet implemented, clause ignored}} - // expected-warning@+3{{OpenACC clause 'independent' not yet implemented, clause ignored}} - // expected-warning@+2{{OpenACC clause 'auto' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+3{{OpenACC clause 'seq' not yet implemented, clause ignored}} + // expected-warning@+2{{OpenACC clause 'independent' not yet implemented, clause ignored}} + // expected-warning@+1{{OpenACC clause 'auto' not yet implemented, clause ignored}} #pragma acc loop seq independent, auto + for(;;){} // expected-warning@+4{{OpenACC clause 'seq' not yet implemented, clause ignored}} // expected-warning@+3{{OpenACC clause 'independent' not yet implemented, clause ignored}} @@ -67,65 +67,57 @@ void func() { // expected-warning@+2{{OpenACC clause 'auto' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC construct 'serial loop' not yet implemented, pragma ignored}} #pragma acc serial loop seq, independent auto - {} + for(;;){} // expected-warning@+4{{OpenACC clause 'seq' not yet implemented, clause ignored}} // expected-warning@+3{{OpenACC clause 'independent' not yet implemented, clause ignored}} // expected-warning@+2{{OpenACC clause 'auto' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC construct 'parallel loop' not yet implemented, pragma ignored}} #pragma acc parallel loop seq independent, auto - {} + for(;;){} - // expected-error@+2{{expected identifier}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+1{{expected identifier}} #pragma acc loop , seq + for(;;){} - // expected-error@+3{{expected identifier}} - // expected-warning@+2{{OpenACC clause 'seq' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected identifier}} + // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc loop seq, + for(;;){} - // expected-error@+2{{expected '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+1{{expected '('}} #pragma acc loop collapse for(;;){} - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+1{{expected expression}} #pragma acc loop collapse() for(;;){} - // expected-error@+3{{invalid tag 'unknown' on 'collapse' clause}} - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{invalid tag 'unknown' on 'collapse' clause}} + // expected-error@+1{{expected expression}} #pragma acc loop collapse(unknown:) for(;;){} - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+1{{expected expression}} #pragma acc loop collapse(force:) for(;;){} - // expected-error@+3{{invalid tag 'unknown' on 'collapse' clause}} - // expected-warning@+2{{OpenACC clause 'collapse' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{invalid tag 'unknown' on 'collapse' clause}} + // expected-warning@+1{{OpenACC clause 'collapse' not yet implemented, clause ignored}} #pragma acc loop collapse(unknown:5) for(;;){} - // expected-warning@+2{{OpenACC clause 'collapse' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'collapse' not yet implemented, clause ignored}} #pragma acc loop collapse(force:5) for(;;){} - // expected-warning@+2{{OpenACC clause 'collapse' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'collapse' not yet implemented, clause ignored}} #pragma acc loop collapse(5) for(;;){} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop collapse(5, 6) for(;;){} } @@ -989,108 +981,108 @@ void IntExprParsing() { #pragma acc set default_async(returns_int()) - // expected-warning@+2{{OpenACC clause 'vector' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} #pragma acc loop vector - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+1{{expected expression}} #pragma acc loop vector() - // expected-error@+3{{invalid tag 'invalid' on 'vector' clause}} - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'invalid' on 'vector' clause}} + // expected-error@+1{{expected expression}} #pragma acc loop vector(invalid:) - // expected-error@+3{{invalid tag 'invalid' on 'vector' clause}} - // expected-warning@+2{{OpenACC clause 'vector' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'invalid' on 'vector' clause}} + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} #pragma acc loop vector(invalid:5) - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+1{{expected expression}} #pragma acc loop vector(length:) - // expected-error@+3{{invalid tag 'num' on 'vector' clause}} - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'num' on 'vector' clause}} + // expected-error@+1{{expected expression}} #pragma acc loop vector(num:) - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop vector(5, 4) - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop vector(length:6,4) - // expected-error@+4{{invalid tag 'num' on 'vector' clause}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+3{{invalid tag 'num' on 'vector' clause}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop vector(num:6,4) - // expected-warning@+2{{OpenACC clause 'vector' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} #pragma acc loop vector(5) - // expected-error@+3{{invalid tag 'num' on 'vector' clause}} - // expected-warning@+2{{OpenACC clause 'vector' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'num' on 'vector' clause}} + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} #pragma acc loop vector(num:5) - // expected-warning@+2{{OpenACC clause 'vector' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} #pragma acc loop vector(length:5) - // expected-warning@+2{{OpenACC clause 'vector' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} #pragma acc loop vector(returns_int()) - // expected-warning@+2{{OpenACC clause 'vector' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} #pragma acc loop vector(length:returns_int()) + for(;;); - // expected-warning@+2{{OpenACC clause 'worker' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'worker' not yet implemented, clause ignored}} #pragma acc loop worker - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+1{{expected expression}} #pragma acc loop worker() - // expected-error@+3{{invalid tag 'invalid' on 'worker' clause}} - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'invalid' on 'worker' clause}} + // expected-error@+1{{expected expression}} #pragma acc loop worker(invalid:) - // expected-error@+3{{invalid tag 'invalid' on 'worker' clause}} - // expected-warning@+2{{OpenACC clause 'worker' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'invalid' on 'worker' clause}} + // expected-warning@+1{{OpenACC clause 'worker' not yet implemented, clause ignored}} #pragma acc loop worker(invalid:5) - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+1{{expected expression}} #pragma acc loop worker(num:) - // expected-error@+3{{invalid tag 'length' on 'worker' clause}} - // expected-error@+2{{expected expression}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'length' on 'worker' clause}} + // expected-error@+1{{expected expression}} #pragma acc loop worker(length:) - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop worker(5, 4) - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop worker(num:6,4) - // expected-error@+4{{invalid tag 'length' on 'worker' clause}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+3{{invalid tag 'length' on 'worker' clause}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop worker(length:6,4) - // expected-warning@+2{{OpenACC clause 'worker' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-warning@+1{{OpenACC clause 'worker' not yet implemented, clause ignored}} #pragma acc loop worker(5) - // expected-error@+3{{invalid tag 'length' on 'worker' clause}} - // expected-warning@+2{{OpenACC clause 'worker' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'length' on 'worker' clause}} + // expected-warning@+1{{OpenACC clause 'worker' not yet implemented, clause ignored}} #pragma acc loop worker(length:5) - // expected-warning@+2{{OpenACC clause 'worker' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-warning@+1{{OpenACC clause 'worker' not yet implemented, clause ignored}} #pragma acc loop worker(num:5) - // expected-warning@+2{{OpenACC clause 'worker' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-warning@+1{{OpenACC clause 'worker' not yet implemented, clause ignored}} #pragma acc loop worker(returns_int()) - // expected-error@+3{{invalid tag 'length' on 'worker' clause}} - // expected-warning@+2{{OpenACC clause 'worker' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + for(;;); + // expected-error@+2{{invalid tag 'length' on 'worker' clause}} + // expected-warning@+1{{OpenACC clause 'worker' not yet implemented, clause ignored}} #pragma acc loop worker(length:returns_int()) + for(;;); } void device_type() { @@ -1236,238 +1228,196 @@ void AsyncArgument() { void Tile() { int* Foo; - // expected-error@+2{{expected '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+1{{expected '('}} #pragma acc loop tile for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop tile( for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile() for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop tile(, for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(,) for(;;){} - // expected-error@+3{{use of undeclared identifier 'invalid'}} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{use of undeclared identifier 'invalid'}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(returns_int(), *, invalid, *) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(returns_int() *, Foo, *) for(;;){} - // expected-error@+3{{indirection requires pointer operand ('int' invalid)}} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{indirection requires pointer operand ('int' invalid)}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(* returns_int() , *) for(;;){} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(*) for(;;){} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(*Foo, *Foo) for(;;){} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(5) for(;;){} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(*, 5) for(;;){} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(5, *) for(;;){} - // expected-warning@+2{{OpenACC clause 'tile' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'tile' not yet implemented, clause ignored}} #pragma acc loop tile(5, *, 3, *) for(;;){} } void Gang() { - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang( for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang() for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(5, *) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(*) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(5, num:*) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(num:5, *) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(num:5, num:*) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(num:*) for(;;){} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(dim:5) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(dim:5, dim:*) for(;;){} - // expected-error@+3{{expected expression}} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(dim:*) for(;;){} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(static:*) for(;;){} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(static:*, static:5) for(;;){} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(static:*, 5) for(;;){} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(static:45, 5) for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(static:45, for(;;){} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(static:45 for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(static:*, for(;;){} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(static:* for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(45, for(;;){} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(45 for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(num:45, for(;;){} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(num:45 for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+3{{expected expression}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(dim:45, for(;;){} - // expected-error@+3{{expected ')'}} - // expected-note@+2{{to match this '('}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} #pragma acc loop gang(dim:45 for(;;){} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(static:*, dim:returns_int(), 5) for(;;){} - // expected-warning@+2{{OpenACC clause 'gang' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'gang' not yet implemented, clause ignored}} #pragma acc loop gang(num: 32, static:*, dim:returns_int(), 5) for(;;){} diff --git a/clang/test/ParserOpenACC/parse-clauses.cpp b/clang/test/ParserOpenACC/parse-clauses.cpp index 702eb75ca8902..b7e252e892bea 100644 --- a/clang/test/ParserOpenACC/parse-clauses.cpp +++ b/clang/test/ParserOpenACC/parse-clauses.cpp @@ -2,13 +2,11 @@ template<unsigned I, typename T> void templ() { - // expected-warning@+2{{OpenACC clause 'collapse' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'collapse' not yet implemented, clause ignored}} #pragma acc loop collapse(I) for(;;){} - // expected-warning@+2{{OpenACC clause 'collapse' not yet implemented, clause ignored}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-warning@+1{{OpenACC clause 'collapse' not yet implemented, clause ignored}} #pragma acc loop collapse(T::value) for(;;){} diff --git a/clang/test/ParserOpenACC/parse-constructs.c b/clang/test/ParserOpenACC/parse-constructs.c index ecedfd9e9e6d6..ea75360cc1351 100644 --- a/clang/test/ParserOpenACC/parse-constructs.c +++ b/clang/test/ParserOpenACC/parse-constructs.c @@ -82,8 +82,7 @@ void func() { // expected-warning@+1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}} #pragma acc host_data clause list for(;;){} - // expected-error@+2{{invalid OpenACC clause 'clause'}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} + // expected-error@+1{{invalid OpenACC clause 'clause'}} #pragma acc loop clause list for(;;){} // expected-error@+1{{invalid OpenACC clause 'invalid'}} diff --git a/clang/test/SemaOpenACC/compute-construct-async-clause.c b/clang/test/SemaOpenACC/compute-construct-async-clause.c index 999db74ffbb8b..fe41c5d0897a4 100644 --- a/clang/test/SemaOpenACC/compute-construct-async-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-async-clause.c @@ -39,8 +39,7 @@ void Test() { #pragma acc kernels async(SomeE) while(1); - // expected-error@+2{{OpenACC 'async' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'async' clause is not valid on 'loop' directive}} #pragma acc loop async(1) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-attach-clause.c b/clang/test/SemaOpenACC/compute-construct-attach-clause.c index 7696620271818..1d204094de12a 100644 --- a/clang/test/SemaOpenACC/compute-construct-attach-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-attach-clause.c @@ -59,8 +59,7 @@ void uses() { #pragma acc parallel attach(s.PtrMem) while (1); - // expected-error@+2{{OpenACC 'attach' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} #pragma acc loop attach(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-copy-clause.c b/clang/test/SemaOpenACC/compute-construct-copy-clause.c index 7adf0e18fa042..284813f213529 100644 --- a/clang/test/SemaOpenACC/compute-construct-copy-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-copy-clause.c @@ -60,16 +60,13 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc parallel copy((float)ArrayParam[2]) while(1); - // expected-error@+2{{OpenACC 'copy' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'copy' clause is not valid on 'loop' directive}} #pragma acc loop copy(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'pcopy' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'pcopy' clause is not valid on 'loop' directive}} #pragma acc loop pcopy(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop present_or_copy(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-copyin-clause.c b/clang/test/SemaOpenACC/compute-construct-copyin-clause.c index d557357756568..d4dda1e16737c 100644 --- a/clang/test/SemaOpenACC/compute-construct-copyin-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-copyin-clause.c @@ -66,16 +66,13 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc parallel copyin(invalid:(float)ArrayParam[2]) while(1); - // expected-error@+2{{OpenACC 'copyin' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'copyin' clause is not valid on 'loop' directive}} #pragma acc loop copyin(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'pcopyin' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'pcopyin' clause is not valid on 'loop' directive}} #pragma acc loop pcopyin(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'present_or_copyin' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'present_or_copyin' clause is not valid on 'loop' directive}} #pragma acc loop present_or_copyin(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-copyout-clause.c b/clang/test/SemaOpenACC/compute-construct-copyout-clause.c index 432823b6746a3..5692ab0f5660c 100644 --- a/clang/test/SemaOpenACC/compute-construct-copyout-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-copyout-clause.c @@ -66,16 +66,13 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc parallel copyout(invalid:(float)ArrayParam[2]) while(1); - // expected-error@+2{{OpenACC 'copyout' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'copyout' clause is not valid on 'loop' directive}} #pragma acc loop copyout(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'pcopyout' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'pcopyout' clause is not valid on 'loop' directive}} #pragma acc loop pcopyout(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'present_or_copyout' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'present_or_copyout' clause is not valid on 'loop' directive}} #pragma acc loop present_or_copyout(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-create-clause.c b/clang/test/SemaOpenACC/compute-construct-create-clause.c index 319025c9628cf..6ef9551d759ee 100644 --- a/clang/test/SemaOpenACC/compute-construct-create-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-create-clause.c @@ -67,16 +67,13 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc parallel create(invalid:(float)ArrayParam[2]) while(1); - // expected-error@+2{{OpenACC 'create' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'create' clause is not valid on 'loop' directive}} #pragma acc loop create(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'pcreate' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'pcreate' clause is not valid on 'loop' directive}} #pragma acc loop pcreate(LocalInt) for(;;); - // expected-error@+2{{OpenACC 'present_or_create' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'present_or_create' clause is not valid on 'loop' directive}} #pragma acc loop present_or_create(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-default-clause.c b/clang/test/SemaOpenACC/compute-construct-default-clause.c index bcafb02cb4df1..93e8f7c2a6b18 100644 --- a/clang/test/SemaOpenACC/compute-construct-default-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-default-clause.c @@ -43,18 +43,16 @@ void SingleOnly() { #pragma acc data default(none) while(0); - // expected-warning@+2{{OpenACC construct 'loop' not yet implemented}} // expected-error@+1{{OpenACC 'default' clause is not valid on 'loop' directive}} #pragma acc loop default(none) - while(0); + for(;;); // expected-warning@+2{{OpenACC construct 'wait' not yet implemented}} // expected-error@+1{{OpenACC 'default' clause is not valid on 'wait' directive}} #pragma acc wait default(none) while(0); - // expected-error@+2{{OpenACC 'default' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'default' clause is not valid on 'loop' directive}} #pragma acc loop default(present) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-deviceptr-clause.c b/clang/test/SemaOpenACC/compute-construct-deviceptr-clause.c index 8ec911f6dbf1d..44c4cc4e5ec27 100644 --- a/clang/test/SemaOpenACC/compute-construct-deviceptr-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-deviceptr-clause.c @@ -59,8 +59,7 @@ void uses() { #pragma acc parallel deviceptr(s.PtrMem) while (1); - // expected-error@+2{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}} #pragma acc loop deviceptr(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-firstprivate-clause.c b/clang/test/SemaOpenACC/compute-construct-firstprivate-clause.c index 14f5af60cc855..0c26a0b4c9b95 100644 --- a/clang/test/SemaOpenACC/compute-construct-firstprivate-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-firstprivate-clause.c @@ -53,8 +53,7 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc parallel firstprivate((float)ArrayParam[2]) while(1); - // expected-error@+2{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}} #pragma acc loop firstprivate(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-if-clause.c b/clang/test/SemaOpenACC/compute-construct-if-clause.c index 21e7ce413e908..4629b1b2c2bd0 100644 --- a/clang/test/SemaOpenACC/compute-construct-if-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-if-clause.c @@ -60,8 +60,7 @@ void BoolExpr(int *I, float *F) { #pragma acc kernels loop if (*I < *F) while(0); - // expected-error@+2{{OpenACC 'if' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'if' clause is not valid on 'loop' directive}} #pragma acc loop if(I) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-no_create-clause.c b/clang/test/SemaOpenACC/compute-construct-no_create-clause.c index 5afd644462147..6db7d0cca8c32 100644 --- a/clang/test/SemaOpenACC/compute-construct-no_create-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-no_create-clause.c @@ -52,8 +52,7 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc parallel no_create((float)ArrayParam[2]) while(1); - // expected-error@+2{{OpenACC 'no_create' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'no_create' clause is not valid on 'loop' directive}} #pragma acc loop no_create(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c b/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c index 9c2a5a781059e..0a86dee4da041 100644 --- a/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c @@ -52,8 +52,7 @@ void Test() { #pragma acc parallel num_gangs(getS(), 1, getS(), 1) while(1); - // expected-error@+2{{OpenACC 'num_gangs' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'num_gangs' clause is not valid on 'loop' directive}} #pragma acc loop num_gangs(1) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c b/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c index a84bd3699536a..808609cf2a0fb 100644 --- a/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c @@ -31,8 +31,7 @@ void Test() { #pragma acc kernels num_workers(SomeE) while(1); - // expected-error@+2{{OpenACC 'num_workers' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'num_workers' clause is not valid on 'loop' directive}} #pragma acc loop num_workers(1) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-present-clause.c b/clang/test/SemaOpenACC/compute-construct-present-clause.c index 5ace750da7efe..eea2c77657c8d 100644 --- a/clang/test/SemaOpenACC/compute-construct-present-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-present-clause.c @@ -52,8 +52,7 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc parallel present((float)ArrayParam[2]) while(1); - // expected-error@+2{{OpenACC 'present' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'present' clause is not valid on 'loop' directive}} #pragma acc loop present(LocalInt) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-self-clause.c b/clang/test/SemaOpenACC/compute-construct-self-clause.c index 634a2d8857b7e..c79e7e5d3db6d 100644 --- a/clang/test/SemaOpenACC/compute-construct-self-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-self-clause.c @@ -80,8 +80,7 @@ void WarnMaybeNotUsed(int val1, int val2) { #pragma acc parallel if(invalid) self(val1) while(0); - // expected-error@+2{{OpenACC 'self' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'self' clause is not valid on 'loop' directive}} #pragma acc loop self for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c index 83055f81fbb2c..eda2d5e251b25 100644 --- a/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c @@ -31,8 +31,7 @@ void Test() { #pragma acc kernels vector_length(SomeE) while(1); - // expected-error@+2{{OpenACC 'vector_length' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'vector_length' clause is not valid on 'loop' directive}} #pragma acc loop vector_length(1) for(;;); } diff --git a/clang/test/SemaOpenACC/compute-construct-wait-clause.c b/clang/test/SemaOpenACC/compute-construct-wait-clause.c index 0878288ca4a2c..0d0ab52c31dcc 100644 --- a/clang/test/SemaOpenACC/compute-construct-wait-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-wait-clause.c @@ -36,8 +36,7 @@ void uses() { #pragma acc parallel wait(devnum:arr : queues: arr, NC, 5) while(1); - // expected-error@+2{{OpenACC 'wait' clause is not valid on 'loop' directive}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented}} + // expected-error@+1{{OpenACC 'wait' clause is not valid on 'loop' directive}} #pragma acc loop wait for(;;); } diff --git a/clang/test/SemaOpenACC/loop-ast.cpp b/clang/test/SemaOpenACC/loop-ast.cpp new file mode 100644 index 0000000000000..292044f94267b --- /dev/null +++ b/clang/test/SemaOpenACC/loop-ast.cpp @@ -0,0 +1,182 @@ + +// 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 + +#pragma acc loop + for(;;); + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} <orphan> + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: NullStmt + + int array[5]; + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl +#pragma acc loop + for(auto x : array){} + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} <orphan> + // CHECK-NEXT: CXXForRangeStmt + // CHECK: CompoundStmt + +#pragma acc parallel + // CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel + // CHECK-NEXT: CompoundStmt + { +#pragma acc parallel + // CHECK-NEXT: OpenACCComputeConstruct [[PAR_ADDR:[0-9a-fx]+]] {{.*}}parallel + // CHECK-NEXT: CompoundStmt + { +#pragma acc loop + for(;;); + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR]] + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: NullStmt + } + } +} + +template<typename T> +void TemplFunc() { + // CHECK-LABEL: FunctionTemplateDecl {{.*}}TemplFunc + // CHECK-NEXT: TemplateTypeParmDecl + // CHECK-NEXT: FunctionDecl{{.*}}TemplFunc + // CHECK-NEXT: CompoundStmt + +#pragma acc loop + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} <orphan> + for(typename T::type t = 0; t < 5;++t) { + // CHECK-NEXT: ForStmt + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl{{.*}} referenced t 'typename T::type' + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 0 + // CHECK-NEXT: <<<NULL>> + // CHECK-NEXT: BinaryOperator{{.*}} '<dependent type>' '<' + // CHECK-NEXT: DeclRefExpr {{.*}} 'typename T::type' lvalue Var + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 5 + // CHECK-NEXT: UnaryOperator{{.*}} '<dependent type>' lvalue prefix '++' + // CHECK-NEXT: DeclRefExpr {{.*}} 'typename T::type' lvalue Var + // CHECK-NEXT: CompoundStmt + typename T::type I; + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl{{.*}} I 'typename T::type' + + } + +#pragma acc parallel + { + // CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel + // CHECK-NEXT: CompoundStmt +#pragma acc parallel + { + // CHECK-NEXT: OpenACCComputeConstruct [[PAR_ADDR_UNINST:[0-9a-fx]+]] {{.*}}parallel + // CHECK-NEXT: CompoundStmt +#pragma acc loop + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_UNINST]] + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: NullStmt + for(;;); + +#pragma acc loop + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_UNINST]] + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: NullStmt + for(;;); + } + } + + typename T::type array[5]; + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl + +#pragma acc loop + for(auto x : array){} + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} <orphan> + // CHECK-NEXT: CXXForRangeStmt + // CHECK: CompoundStmt + + // Instantiation: + // CHECK-NEXT: FunctionDecl{{.*}} TemplFunc 'void ()' implicit_instantiation + // CHECK-NEXT: TemplateArgument type 'S' + // CHECK-NEXT: RecordType{{.*}} 'S' + // CHECK-NEXT: CXXRecord{{.*}} 'S' + // CHECK-NEXT: CompoundStmt + + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} <orphan> + // CHECK-NEXT: ForStmt + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl{{.*}} used t 'typename S::type':'int' + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 0 + // CHECK-NEXT: <<<NULL>> + // CHECK-NEXT: BinaryOperator{{.*}} 'bool' '<' + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'typename S::type':'int' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr {{.*}} 'typename S::type':'int' lvalue Var + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 5 + // CHECK-NEXT: UnaryOperator{{.*}} 'typename S::type':'int' lvalue prefix '++' + // CHECK-NEXT: DeclRefExpr {{.*}} 'typename S::type':'int' lvalue Var + // CHECK-NEXT: CompoundStmt + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl{{.*}} I 'typename S::type':'int' + + // CHECK-NEXT: OpenACCComputeConstruct {{.*}}parallel + // CHECK-NEXT: CompoundStmt + // + // CHECK-NEXT: OpenACCComputeConstruct [[PAR_ADDR_INST:[0-9a-fx]+]] {{.*}}parallel + // CHECK-NEXT: CompoundStmt + + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_INST]] + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: NullStmt + + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_INST]] + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: <<<NULL>>> + // CHECK-NEXT: NullStmt + + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} <orphan> + // CHECK-NEXT: CXXForRangeStmt + // CHECK: CompoundStmt +} + +struct S { + using type = int; +}; + +void use() { + TemplFunc<S>(); +} +#endif + diff --git a/clang/test/SemaOpenACC/loop-loc-and-stmt.c b/clang/test/SemaOpenACC/loop-loc-and-stmt.c new file mode 100644 index 0000000000000..36c6743f9843b --- /dev/null +++ b/clang/test/SemaOpenACC/loop-loc-and-stmt.c @@ -0,0 +1,38 @@ +// RUN: %clang_cc1 %s -verify -fopenacc + +// expected-error@+1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}} +#pragma acc loop + +// expected-error@+1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}} +#pragma acc loop +int foo; + +struct S { +// expected-error@+1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}} +#pragma acc loop + int i; +}; + +void func() { + // expected-error@+2{{expected expression}} +#pragma acc loop + int foo; + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + while(0); + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + do{}while(0); + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + {} + +#pragma acc loop + for(;;); +} diff --git a/clang/test/SemaOpenACC/loop-loc-and-stmt.cpp b/clang/test/SemaOpenACC/loop-loc-and-stmt.cpp new file mode 100644 index 0000000000000..5d50145b7c882 --- /dev/null +++ b/clang/test/SemaOpenACC/loop-loc-and-stmt.cpp @@ -0,0 +1,80 @@ +// RUN: %clang_cc1 %s -verify -fopenacc +// +// expected-error@+1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}} +#pragma acc loop + +// expected-error@+1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}} +#pragma acc loop +int foo; + +struct S { +// expected-error@+1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}} +#pragma acc loop + int i; + + void mem_func() { + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + int foo; + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + while(0); + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + do{}while(0); + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + {} + +#pragma acc loop + for(;;); + + int array[5]; + +#pragma acc loop + for(auto X : array){} +} +}; + +template<typename T> +void templ_func() { + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + int foo; + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + while(T{}); + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + do{}while(0); + + // expected-error@+3{{OpenACC 'loop' construct can only be applied to a 'for' loop}} + // expected-note@+1{{'loop' construct is here}} +#pragma acc loop + {} + +#pragma acc loop + for(T i;;); + + T array[5]; + +#pragma acc loop + for(auto X : array){} +} + +void use() { + templ_func<int>(); +} + diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp index 49ed60d990ca6..916e941cfbde1 100644 --- a/clang/tools/libclang/CIndex.cpp +++ b/clang/tools/libclang/CIndex.cpp @@ -2170,6 +2170,7 @@ class EnqueueVisitor : public ConstStmtVisitor<EnqueueVisitor, void>, void VisitRequiresExpr(const RequiresExpr *E); void VisitCXXParenListInitExpr(const CXXParenListInitExpr *E); void VisitOpenACCComputeConstruct(const OpenACCComputeConstruct *D); + void VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *D); void VisitOMPExecutableDirective(const OMPExecutableDirective *D); void VisitOMPLoopBasedDirective(const OMPLoopBasedDirective *D); void VisitOMPLoopDirective(const OMPLoopDirective *D); @@ -3496,6 +3497,12 @@ void EnqueueVisitor::VisitOpenACCComputeConstruct( EnqueueChildren(Clause); } +void EnqueueVisitor::VisitOpenACCLoopConstruct(const OpenACCLoopConstruct *C) { + EnqueueChildren(C); + for (auto *Clause : C->clauses()) + EnqueueChildren(Clause); +} + void EnqueueVisitor::VisitAnnotateAttr(const AnnotateAttr *A) { EnqueueChildren(A); } @@ -6234,6 +6241,8 @@ CXString clang_getCursorKindSpelling(enum CXCursorKind Kind) { return cxstring::createRef("ConceptDecl"); case CXCursor_OpenACCComputeConstruct: return cxstring::createRef("OpenACCComputeConstruct"); + case CXCursor_OpenACCLoopConstruct: + return cxstring::createRef("OpenACCLoopConstruct"); } llvm_unreachable("Unhandled CXCursorKind"); diff --git a/clang/tools/libclang/CXCursor.cpp b/clang/tools/libclang/CXCursor.cpp index 9325a16d2a848..38002052227cd 100644 --- a/clang/tools/libclang/CXCursor.cpp +++ b/clang/tools/libclang/CXCursor.cpp @@ -873,6 +873,9 @@ CXCursor cxcursor::MakeCXCursor(const Stmt *S, const Decl *Parent, case Stmt::OpenACCComputeConstructClass: K = CXCursor_OpenACCComputeConstruct; break; + case Stmt::OpenACCLoopConstructClass: + K = CXCursor_OpenACCLoopConstruct; + 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
