[clang] [OpenACC] Loop construct basic Sema and AST work (PR #93742)
via cfe-commits
cfe-commits at lists.llvm.org
Wed May 29 14:24:39 PDT 2024
llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT-->
@llvm/pr-subscribers-clang-codegen
@llvm/pr-subscribers-clang
Author: Erich Keane (erichkeane)
<details>
<summary>Changes</summary>
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).
---
Patch is 74.83 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/93742.diff
31 Files Affected:
- (modified) clang/include/clang-c/Index.h (+5-1)
- (modified) clang/include/clang/AST/RecursiveASTVisitor.h (+2)
- (modified) clang/include/clang/AST/StmtOpenACC.h (+70-1)
- (modified) clang/include/clang/AST/TextNodeDumper.h (+1)
- (modified) clang/include/clang/Basic/DiagnosticSemaKinds.td (+3)
- (modified) clang/include/clang/Basic/StmtNodes.td (+1)
- (modified) clang/include/clang/Sema/SemaOpenACC.h (+26-1)
- (modified) clang/include/clang/Serialization/ASTBitCodes.h (+1)
- (modified) clang/lib/AST/StmtOpenACC.cpp (+90-1)
- (modified) clang/lib/AST/StmtPrinter.cpp (+13)
- (modified) clang/lib/AST/StmtProfile.cpp (+8)
- (modified) clang/lib/AST/TextNodeDumper.cpp (+7)
- (modified) clang/lib/CodeGen/CGStmt.cpp (+3)
- (modified) clang/lib/CodeGen/CodeGenFunction.h (+7)
- (modified) clang/lib/Parse/ParseOpenACC.cpp (+5-3)
- (modified) clang/lib/Sema/SemaExceptionSpec.cpp (+1)
- (modified) clang/lib/Sema/SemaOpenACC.cpp (+56-4)
- (modified) clang/lib/Sema/TreeTransform.h (+39-5)
- (modified) clang/lib/Serialization/ASTReaderStmt.cpp (+11)
- (modified) clang/lib/Serialization/ASTWriterStmt.cpp (+6)
- (modified) clang/lib/StaticAnalyzer/Core/ExprEngine.cpp (+1)
- (added) clang/test/AST/ast-print-openacc-loop-construct.cpp (+9)
- (modified) clang/test/ParserOpenACC/parse-clauses.c (+180-230)
- (modified) clang/test/ParserOpenACC/parse-clauses.cpp (+2-4)
- (modified) clang/test/ParserOpenACC/parse-constructs.c (+1-2)
- (modified) clang/test/SemaOpenACC/compute-construct-default-clause.c (+1-2)
- (added) clang/test/SemaOpenACC/loop-ast.cpp (+164)
- (added) clang/test/SemaOpenACC/loop-loc-and-stmt.c (+38)
- (added) clang/test/SemaOpenACC/loop-loc-and-stmt.cpp (+80)
- (modified) clang/tools/libclang/CIndex.cpp (+9)
- (modified) clang/tools/libclang/CXCursor.cpp (+3)
``````````diff
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 4bbb4380cdd7f..bb6ff334dd32d 100644
--- a/clang/include/clang/AST/RecursiveASTVisitor.h
+++ b/clang/include/clang/AST/RecursiveASTVisitor.h
@@ -3996,6 +3996,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 63fa16c9ec47c..de82b438b0eb0 100644
--- a/clang/include/clang/AST/TextNodeDumper.h
+++ b/clang/include/clang/AST/TextNodeDumper.h
@@ -407,6 +407,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 f15cba63624ea..b3d985c5ff48c 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12406,6 +12406,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..e235449eb9959 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 VisitOpenACCComputeConstruct(OpenACCComputeConstruct *C) {
+ // Stop searching if we find a compute construct.
+ return false;
+ }
+ bool VisitOpenACCLoopConstruct(OpenACCLoopConstruct *C) {
+ // Stop searching if we find a loop construct, after taking ownership of
+ // it.
+ C->setParentComputeConstruct(Construct);
+ return false;
+ }
+ };
+
+ 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 a0eedc71ea220..194370c1d82e0 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -2848,3 +2848,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::OpenACCComputeConstr...
[truncated]
``````````
</details>
https://github.com/llvm/llvm-project/pull/93742
More information about the cfe-commits
mailing list