diff --git a/clang/include/clang-c/Index.h b/clang/include/clang-c/Index.h index 365b607c7411..ce2282937f86 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 99093aa17972..aa55e2e7e871 100644 --- a/clang/include/clang/AST/RecursiveASTVisitor.h +++ b/clang/include/clang/AST/RecursiveASTVisitor.h @@ -4000,6 +4000,8 @@ bool RecursiveASTVisitor::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 04daf511f587..b3aea09be03d 100644 --- a/clang/include/clang/AST/StmtOpenACC.h +++ b/clang/include/clang/AST/StmtOpenACC.h @@ -113,6 +113,8 @@ public: return const_cast(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 @@ public: static OpenACCComputeConstruct * Create(const ASTContext &C, OpenACCDirectiveKind K, SourceLocation BeginLoc, SourceLocation DirectiveLoc, SourceLocation EndLoc, - ArrayRef Clauses, Stmt *StructuredBlock); + ArrayRef Clauses, Stmt *StructuredBlock, + ArrayRef AssociatedLoopConstructs); Stmt *getStructuredBlock() { return getAssociatedStmt(); } const Stmt *getStructuredBlock() const { return const_cast(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 { + // 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 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 Clauses, + Stmt *Loop); + + Stmt *getLoop() { return getAssociatedStmt(); } + const Stmt *getLoop() const { + return const_cast(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 caa33abd99e4..abfafcaef271 100644 --- a/clang/include/clang/AST/TextNodeDumper.h +++ b/clang/include/clang/AST/TextNodeDumper.h @@ -408,6 +408,7 @@ public: 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 e34eb692941b..50332966a7e3 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 305f19daa4a9..6ca08abdb14f 100644 --- a/clang/include/clang/Basic/StmtNodes.td +++ b/clang/include/clang/Basic/StmtNodes.td @@ -302,3 +302,4 @@ def OpenACCConstructStmt : StmtNode; def OpenACCAssociatedStmtConstruct : StmtNode; def OpenACCComputeConstruct : StmtNode; +def OpenACCLoopConstruct : StmtNode; diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h index 66144de4340a..a5f2a8bf7465 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 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; @@ -394,7 +404,8 @@ public: 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 @@ public: 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 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 fe1bd47348be..f59ff6af4c76 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 47899b344c97..2d864a288579 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 Clauses, Stmt *StructuredBlock) { + ArrayRef Clauses, Stmt *StructuredBlock, + ArrayRef AssociatedLoopConstructs) { void *Mem = C.Allocate( OpenACCComputeConstruct::totalSizeToAlloc( 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 { + 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(), + getTrailingObjects() + NumClauses); + setClauseList( + MutableArrayRef(getTrailingObjects(), NumClauses)); +} + +OpenACCLoopConstruct::OpenACCLoopConstruct( + SourceLocation Start, SourceLocation DirLoc, SourceLocation End, + ArrayRef 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(Loop)) && + "Associated Loop not a for loop?"); + // Initialize the trailing storage. + std::uninitialized_copy(Clauses.begin(), Clauses.end(), + getTrailingObjects()); + + setClauseList(MutableArrayRef(getTrailingObjects(), + Clauses.size())); +} + +void OpenACCLoopConstruct::setLoop(Stmt *Loop) { + assert((isa(Loop)) && + "Associated Loop not a for loop?"); + setAssociatedStmt(Loop); +} + +OpenACCLoopConstruct *OpenACCLoopConstruct::CreateEmpty(const ASTContext &C, + unsigned NumClauses) { + void *Mem = + C.Allocate(OpenACCLoopConstruct::totalSizeToAlloc( + NumClauses)); + auto *Inst = new (Mem) OpenACCLoopConstruct(NumClauses); + return Inst; +} + +OpenACCLoopConstruct * +OpenACCLoopConstruct::Create(const ASTContext &C, SourceLocation BeginLoc, + SourceLocation DirLoc, SourceLocation EndLoc, + ArrayRef Clauses, + Stmt *Loop) { + void *Mem = + C.Allocate(OpenACCLoopConstruct::totalSizeToAlloc( + 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 be2d5a2eb6b4..7e030e055126 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 00b8c43af035..6d9a76120cfe 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 0e0e0a86f5cf..b2bf259ec24e 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 << " "; + else + OS << " parent: " << S->getParentComputeConstruct(); +} diff --git a/clang/lib/CodeGen/CGStmt.cpp b/clang/lib/CodeGen/CGStmt.cpp index 99daaa14cf3f..41ac511c52f5 100644 --- a/clang/lib/CodeGen/CGStmt.cpp +++ b/clang/lib/CodeGen/CGStmt.cpp @@ -442,6 +442,9 @@ void CodeGenFunction::EmitStmt(const Stmt *S, ArrayRef Attrs) { case Stmt::OpenACCComputeConstructClass: EmitOpenACCComputeConstruct(cast(*S)); break; + case Stmt::OpenACCLoopConstructClass: + EmitOpenACCLoopConstruct(cast(*S)); + break; } } diff --git a/clang/lib/CodeGen/CodeGenFunction.h b/clang/lib/CodeGen/CodeGenFunction.h index 45585361a4fc..5739fbaaa919 100644 --- a/clang/lib/CodeGen/CodeGenFunction.h +++ b/clang/lib/CodeGen/CodeGenFunction.h @@ -4062,6 +4062,13 @@ public: 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 63afc18783a1..c7b6763b4dbd 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 41bf273d12f2..17acfca6b011 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 5b4d01860c0b..92da218010c9 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 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(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 70603ba6c271..07f995edaf3d 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -4041,6 +4041,15 @@ public: EndLoc, Clauses, StrBlock); } + StmtResult RebuildOpenACCLoopConstruct(SourceLocation BeginLoc, + SourceLocation DirLoc, + SourceLocation EndLoc, + ArrayRef Clauses, + StmtResult Loop) { + return getSema().OpenACC().ActOnEndStmtDirective( + OpenACCDirectiveKind::Loop, BeginLoc, DirLoc, EndLoc, Clauses, Loop); + } + private: TypeLoc TransformTypeInObjectScope(TypeLoc TL, QualType ObjectType, @@ -11541,8 +11550,31 @@ template StmtResult TreeTransform::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())) + return StmtError(); + + llvm::SmallVector 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->getBeginLoc(), C->getDirectiveKind(), StrBlock); + + return getDerived().RebuildOpenACCComputeConstruct( + C->getDirectiveKind(), C->getBeginLoc(), C->getDirectiveLoc(), + C->getEndLoc(), TransformedClauses, StrBlock); +} + +template +StmtResult +TreeTransform::TransformOpenACCLoopConstruct(OpenACCLoopConstruct *C) { + + getSema().OpenACC().ActOnConstruct(C->getDirectiveKind(), C->getBeginLoc()); if (getSema().OpenACC().ActOnStartStmtDirective(C->getDirectiveKind(), C->getBeginLoc())) @@ -11552,14 +11584,16 @@ StmtResult TreeTransform::TransformOpenACCComputeConstruct( getDerived().TransformOpenACCClauseList(C->getDirectiveKind(), C->clauses()); - // Transform Structured Block. - StmtResult StrBlock = getDerived().TransformStmt(C->getStructuredBlock()); - StrBlock = - getSema().OpenACC().ActOnAssociatedStmt(C->getDirectiveKind(), StrBlock); + // 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().RebuildOpenACCComputeConstruct( - C->getDirectiveKind(), C->getBeginLoc(), C->getDirectiveLoc(), - C->getEndLoc(), TransformedClauses, StrBlock); + return getDerived().RebuildOpenACCLoopConstruct( + C->getBeginLoc(), C->getDirectiveLoc(), C->getEndLoc(), + TransformedClauses, Loop); } //===----------------------------------------------------------------------===// diff --git a/clang/lib/Serialization/ASTReaderStmt.cpp b/clang/lib/Serialization/ASTReaderStmt.cpp index bea2b9498910..67ef17025191 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 3c586b270fbf..1a98e30e0f89 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 793f3a63ea29..290d96611d46 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 000000000000..21c92b17317e --- /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 49e749feb2ec..cb118f69fb44 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@+1{{expected identifier}} +#pragma acc loop , seq + for(;;){} + // expected-error@+2{{expected identifier}} - // expected-warning@+1{{OpenACC construct 'loop' not yet implemented, pragma ignored}} -#pragma acc loop , seq - - // 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-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}} -#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}} -#pragma acc loop vector(5) + for(;;); // 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}} + // expected-error@+2{{expected ')'}} + // expected-note@+1{{to match this '('}} +#pragma acc loop vector(num:6,4) + for(;;); + // expected-warning@+1{{OpenACC clause 'vector' not yet implemented, clause ignored}} +#pragma acc loop vector(5) + 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 702eb75ca890..b7e252e892be 100644 --- a/clang/test/ParserOpenACC/parse-clauses.cpp +++ b/clang/test/ParserOpenACC/parse-clauses.cpp @@ -2,13 +2,11 @@ template 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 ecedfd9e9e6d..ea75360cc135 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 999db74ffbb8..fe41c5d0897a 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 769662027181..1d204094de12 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 7adf0e18fa04..284813f21352 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 d55735775656..d4dda1e16737 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 432823b6746a..5692ab0f5660 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 319025c9628c..6ef9551d759e 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 bcafb02cb4df..93e8f7c2a6b1 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 8ec911f6dbf1..44c4cc4e5ec2 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 14f5af60cc85..0c26a0b4c9b9 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 21e7ce413e90..4629b1b2c2bd 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 5afd64446214..6db7d0cca8c3 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 9c2a5a781059..0a86dee4da04 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 a84bd3699536..808609cf2a0f 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 5ace750da7ef..eea2c77657c8 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 634a2d8857b7..c79e7e5d3db6 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 83055f81fbb2..eda2d5e251b2 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 0878288ca4a2..0d0ab52c31dc 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 000000000000..292044f94267 --- /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{{.*}} + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: NullStmt + + int array[5]; + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl +#pragma acc loop + for(auto x : array){} + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} + // 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: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: NullStmt + } + } +} + +template +void TemplFunc() { + // CHECK-LABEL: FunctionTemplateDecl {{.*}}TemplFunc + // CHECK-NEXT: TemplateTypeParmDecl + // CHECK-NEXT: FunctionDecl{{.*}}TemplFunc + // CHECK-NEXT: CompoundStmt + +#pragma acc loop + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} + 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: <<> + // CHECK-NEXT: BinaryOperator{{.*}} '' '<' + // CHECK-NEXT: DeclRefExpr {{.*}} 'typename T::type' lvalue Var + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 5 + // CHECK-NEXT: UnaryOperator{{.*}} '' 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: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: NullStmt + for(;;); + +#pragma acc loop + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_UNINST]] + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // 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{{.*}} + // 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{{.*}} + // CHECK-NEXT: ForStmt + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl{{.*}} used t 'typename S::type':'int' + // CHECK-NEXT: IntegerLiteral {{.*}} 'int' 0 + // CHECK-NEXT: <<> + // CHECK-NEXT: BinaryOperator{{.*}} 'bool' '<' + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'typename S::type':'int' + // 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: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: NullStmt + + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} parent: [[PAR_ADDR_INST]] + // CHECK-NEXT: ForStmt + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: <<>> + // CHECK-NEXT: NullStmt + + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl + // CHECK-NEXT: OpenACCLoopConstruct{{.*}} + // CHECK-NEXT: CXXForRangeStmt + // CHECK: CompoundStmt +} + +struct S { + using type = int; +}; + +void use() { + TemplFunc(); +} +#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 000000000000..36c6743f9843 --- /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 000000000000..5d50145b7c88 --- /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 +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(); +} + diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp index 49ed60d990ca..916e941cfbde 100644 --- a/clang/tools/libclang/CIndex.cpp +++ b/clang/tools/libclang/CIndex.cpp @@ -2170,6 +2170,7 @@ public: 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 9325a16d2a84..38002052227c 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;