[clang] 42f4e50 - [OpenACC] Loop construct basic Sema and AST work (#93742)
via cfe-commits
cfe-commits at lists.llvm.org
Wed Jun 5 06:21:51 PDT 2024
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 at +1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}}
#pragma acc host_data if_present, if_present
- // expected-warning at +4{{OpenACC clause 'seq' not yet implemented, clause ignored}}
- // expected-warning at +3{{OpenACC clause 'independent' not yet implemented, clause ignored}}
- // expected-warning at +2{{OpenACC clause 'auto' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +3{{OpenACC clause 'seq' not yet implemented, clause ignored}}
+ // expected-warning at +2{{OpenACC clause 'independent' not yet implemented, clause ignored}}
+ // expected-warning at +1{{OpenACC clause 'auto' not yet implemented, clause ignored}}
#pragma acc loop seq independent auto
+ for(;;){}
- // expected-warning at +4{{OpenACC clause 'seq' not yet implemented, clause ignored}}
- // expected-warning at +3{{OpenACC clause 'independent' not yet implemented, clause ignored}}
- // expected-warning at +2{{OpenACC clause 'auto' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +3{{OpenACC clause 'seq' not yet implemented, clause ignored}}
+ // expected-warning at +2{{OpenACC clause 'independent' not yet implemented, clause ignored}}
+ // expected-warning at +1{{OpenACC clause 'auto' not yet implemented, clause ignored}}
#pragma acc loop seq, independent auto
+ for(;;){}
- // expected-warning at +4{{OpenACC clause 'seq' not yet implemented, clause ignored}}
- // expected-warning at +3{{OpenACC clause 'independent' not yet implemented, clause ignored}}
- // expected-warning at +2{{OpenACC clause 'auto' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +3{{OpenACC clause 'seq' not yet implemented, clause ignored}}
+ // expected-warning at +2{{OpenACC clause 'independent' not yet implemented, clause ignored}}
+ // expected-warning at +1{{OpenACC clause 'auto' not yet implemented, clause ignored}}
#pragma acc loop seq independent, auto
+ for(;;){}
// expected-warning at +4{{OpenACC clause 'seq' not yet implemented, clause ignored}}
// expected-warning at +3{{OpenACC clause 'independent' not yet implemented, clause ignored}}
@@ -67,65 +67,57 @@ void func() {
// expected-warning at +2{{OpenACC clause 'auto' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC construct 'serial loop' not yet implemented, pragma ignored}}
#pragma acc serial loop seq, independent auto
- {}
+ for(;;){}
// expected-warning at +4{{OpenACC clause 'seq' not yet implemented, clause ignored}}
// expected-warning at +3{{OpenACC clause 'independent' not yet implemented, clause ignored}}
// expected-warning at +2{{OpenACC clause 'auto' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC construct 'parallel loop' not yet implemented, pragma ignored}}
#pragma acc parallel loop seq independent, auto
- {}
+ for(;;){}
- // expected-error at +2{{expected identifier}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +1{{expected identifier}}
#pragma acc loop , seq
+ for(;;){}
- // expected-error at +3{{expected identifier}}
- // expected-warning at +2{{OpenACC clause 'seq' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected identifier}}
+ // expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc loop seq,
+ for(;;){}
- // expected-error at +2{{expected '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +1{{expected '('}}
#pragma acc loop collapse
for(;;){}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +1{{expected expression}}
#pragma acc loop collapse()
for(;;){}
- // expected-error at +3{{invalid tag 'unknown' on 'collapse' clause}}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{invalid tag 'unknown' on 'collapse' clause}}
+ // expected-error at +1{{expected expression}}
#pragma acc loop collapse(unknown:)
for(;;){}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +1{{expected expression}}
#pragma acc loop collapse(force:)
for(;;){}
- // expected-error at +3{{invalid tag 'unknown' on 'collapse' clause}}
- // expected-warning at +2{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{invalid tag 'unknown' on 'collapse' clause}}
+ // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
#pragma acc loop collapse(unknown:5)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
#pragma acc loop collapse(force:5)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
#pragma acc loop collapse(5)
for(;;){}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +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 at +2{{OpenACC clause 'vector' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
#pragma acc loop vector
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +1{{expected expression}}
#pragma acc loop vector()
- // expected-error at +3{{invalid tag 'invalid' on 'vector' clause}}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'invalid' on 'vector' clause}}
+ // expected-error at +1{{expected expression}}
#pragma acc loop vector(invalid:)
- // expected-error at +3{{invalid tag 'invalid' on 'vector' clause}}
- // expected-warning at +2{{OpenACC clause 'vector' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'invalid' on 'vector' clause}}
+ // expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
#pragma acc loop vector(invalid:5)
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +1{{expected expression}}
#pragma acc loop vector(length:)
- // expected-error at +3{{invalid tag 'num' on 'vector' clause}}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'num' on 'vector' clause}}
+ // expected-error at +1{{expected expression}}
#pragma acc loop vector(num:)
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop vector(5, 4)
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop vector(length:6,4)
- // expected-error at +4{{invalid tag 'num' on 'vector' clause}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +3{{invalid tag 'num' on 'vector' clause}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop vector(num:6,4)
- // expected-warning at +2{{OpenACC clause 'vector' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
#pragma acc loop vector(5)
- // expected-error at +3{{invalid tag 'num' on 'vector' clause}}
- // expected-warning at +2{{OpenACC clause 'vector' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'num' on 'vector' clause}}
+ // expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
#pragma acc loop vector(num:5)
- // expected-warning at +2{{OpenACC clause 'vector' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
#pragma acc loop vector(length:5)
- // expected-warning at +2{{OpenACC clause 'vector' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
#pragma acc loop vector(returns_int())
- // expected-warning at +2{{OpenACC clause 'vector' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-warning at +1{{OpenACC clause 'vector' not yet implemented, clause ignored}}
#pragma acc loop vector(length:returns_int())
+ for(;;);
- // expected-warning at +2{{OpenACC clause 'worker' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'worker' not yet implemented, clause ignored}}
#pragma acc loop worker
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +1{{expected expression}}
#pragma acc loop worker()
- // expected-error at +3{{invalid tag 'invalid' on 'worker' clause}}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'invalid' on 'worker' clause}}
+ // expected-error at +1{{expected expression}}
#pragma acc loop worker(invalid:)
- // expected-error at +3{{invalid tag 'invalid' on 'worker' clause}}
- // expected-warning at +2{{OpenACC clause 'worker' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'invalid' on 'worker' clause}}
+ // expected-warning at +1{{OpenACC clause 'worker' not yet implemented, clause ignored}}
#pragma acc loop worker(invalid:5)
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +1{{expected expression}}
#pragma acc loop worker(num:)
- // expected-error at +3{{invalid tag 'length' on 'worker' clause}}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'length' on 'worker' clause}}
+ // expected-error at +1{{expected expression}}
#pragma acc loop worker(length:)
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop worker(5, 4)
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop worker(num:6,4)
- // expected-error at +4{{invalid tag 'length' on 'worker' clause}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +3{{invalid tag 'length' on 'worker' clause}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop worker(length:6,4)
- // expected-warning at +2{{OpenACC clause 'worker' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-warning at +1{{OpenACC clause 'worker' not yet implemented, clause ignored}}
#pragma acc loop worker(5)
- // expected-error at +3{{invalid tag 'length' on 'worker' clause}}
- // expected-warning at +2{{OpenACC clause 'worker' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'length' on 'worker' clause}}
+ // expected-warning at +1{{OpenACC clause 'worker' not yet implemented, clause ignored}}
#pragma acc loop worker(length:5)
- // expected-warning at +2{{OpenACC clause 'worker' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-warning at +1{{OpenACC clause 'worker' not yet implemented, clause ignored}}
#pragma acc loop worker(num:5)
- // expected-warning at +2{{OpenACC clause 'worker' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-warning at +1{{OpenACC clause 'worker' not yet implemented, clause ignored}}
#pragma acc loop worker(returns_int())
- // expected-error at +3{{invalid tag 'length' on 'worker' clause}}
- // expected-warning at +2{{OpenACC clause 'worker' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ for(;;);
+ // expected-error at +2{{invalid tag 'length' on 'worker' clause}}
+ // expected-warning at +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 at +2{{expected '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +1{{expected '('}}
#pragma acc loop tile
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop tile(
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile()
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop tile(,
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(,)
for(;;){}
- // expected-error at +3{{use of undeclared identifier 'invalid'}}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{use of undeclared identifier 'invalid'}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(returns_int(), *, invalid, *)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(returns_int() *, Foo, *)
for(;;){}
- // expected-error at +3{{indirection requires pointer operand ('int' invalid)}}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{indirection requires pointer operand ('int' invalid)}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(* returns_int() , *)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(*)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(*Foo, *Foo)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(5)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(*, 5)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(5, *)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'tile' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'tile' not yet implemented, clause ignored}}
#pragma acc loop tile(5, *, 3, *)
for(;;){}
}
void Gang() {
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang()
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(5, *)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(*)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(5, num:*)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(num:5, *)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(num:5, num:*)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(num:*)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(dim:5)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(dim:5, dim:*)
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(dim:*)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(static:*)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(static:*, static:5)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(static:*, 5)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(static:45, 5)
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(static:45,
for(;;){}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(static:45
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(static:*,
for(;;){}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(static:*
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(45,
for(;;){}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(45
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(num:45,
for(;;){}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(num:45
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +3{{expected expression}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(dim:45,
for(;;){}
- // expected-error at +3{{expected ')'}}
- // expected-note at +2{{to match this '('}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +2{{expected ')'}}
+ // expected-note at +1{{to match this '('}}
#pragma acc loop gang(dim:45
for(;;){}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'gang' not yet implemented, clause ignored}}
#pragma acc loop gang(static:*, dim:returns_int(), 5)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'gang' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +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 at +2{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +1{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
#pragma acc loop collapse(I)
for(;;){}
- // expected-warning at +2{{OpenACC clause 'collapse' not yet implemented, clause ignored}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-warning at +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 at +1{{OpenACC construct 'host_data' not yet implemented, pragma ignored}}
#pragma acc host_data clause list
for(;;){}
- // expected-error at +2{{invalid OpenACC clause 'clause'}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented, pragma ignored}}
+ // expected-error at +1{{invalid OpenACC clause 'clause'}}
#pragma acc loop clause list
for(;;){}
// expected-error at +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 at +2{{OpenACC 'async' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'attach' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'copy' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'copy' clause is not valid on 'loop' directive}}
#pragma acc loop copy(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'pcopy' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'pcopy' clause is not valid on 'loop' directive}}
#pragma acc loop pcopy(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'copyin' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'copyin' clause is not valid on 'loop' directive}}
#pragma acc loop copyin(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'pcopyin' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'pcopyin' clause is not valid on 'loop' directive}}
#pragma acc loop pcopyin(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'present_or_copyin' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'copyout' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'copyout' clause is not valid on 'loop' directive}}
#pragma acc loop copyout(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'pcopyout' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'pcopyout' clause is not valid on 'loop' directive}}
#pragma acc loop pcopyout(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'present_or_copyout' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'create' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'create' clause is not valid on 'loop' directive}}
#pragma acc loop create(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'pcreate' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +1{{OpenACC 'pcreate' clause is not valid on 'loop' directive}}
#pragma acc loop pcreate(LocalInt)
for(;;);
- // expected-error at +2{{OpenACC 'present_or_create' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC construct 'loop' not yet implemented}}
// expected-error at +1{{OpenACC 'default' clause is not valid on 'loop' directive}}
#pragma acc loop default(none)
- while(0);
+ for(;;);
// expected-warning at +2{{OpenACC construct 'wait' not yet implemented}}
// expected-error at +1{{OpenACC 'default' clause is not valid on 'wait' directive}}
#pragma acc wait default(none)
while(0);
- // expected-error at +2{{OpenACC 'default' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'if' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'no_create' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'num_gangs' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'num_workers' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'present' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'self' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'vector_length' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +2{{OpenACC 'wait' clause is not valid on 'loop' directive}}
- // expected-warning at +1{{OpenACC construct 'loop' not yet implemented}}
+ // expected-error at +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 at +1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}}
+#pragma acc loop
+
+// expected-error at +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 at +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 at +2{{expected expression}}
+#pragma acc loop
+ int foo;
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ while(0);
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ do{}while(0);
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +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 at +1{{OpenACC construct 'loop' cannot be used here; it can only be used in a statement context}}
+#pragma acc loop
+
+// expected-error at +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 at +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 at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ int foo;
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ while(0);
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ do{}while(0);
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +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 at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ int foo;
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ while(T{});
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +1{{'loop' construct is here}}
+#pragma acc loop
+ do{}while(0);
+
+ // expected-error at +3{{OpenACC 'loop' construct can only be applied to a 'for' loop}}
+ // expected-note at +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;
More information about the cfe-commits
mailing list