[clang] [OpenACC] Private Clause on Compute Constructs (PR #90521)
Erich Keane via cfe-commits
cfe-commits at lists.llvm.org
Tue Apr 30 06:19:36 PDT 2024
https://github.com/erichkeane updated https://github.com/llvm/llvm-project/pull/90521
>From 54ba7d0b7d71d751cf268c3bdfb89bc5ca628a6b Mon Sep 17 00:00:00 2001
From: erichkeane <ekeane at nvidia.com>
Date: Mon, 22 Apr 2024 13:31:52 -0700
Subject: [PATCH 1/2] [OpenACC] Private Clause on Compute Constructs
The private clause is the first that takes a 'var-list', thus this has a
lot of additional work to enable the var-list type. A 'var' is a
traditional variable reference, subscript, member-expression, or array-section,
so checking of these is pretty minor.
Note: This ran into some issues with array-sections (aka sub-arrays)
that will be fixed in a follow-up patch.
---
clang/include/clang/AST/OpenACCClause.h | 85 ++-
.../clang/Basic/DiagnosticSemaKinds.td | 3 +
clang/include/clang/Basic/OpenACCClauses.def | 1 +
clang/include/clang/Parse/Parser.h | 9 +-
clang/include/clang/Sema/SemaOpenACC.h | 36 +-
.../clang/Serialization/ASTRecordReader.h | 3 +
.../clang/Serialization/ASTRecordWriter.h | 3 +
clang/lib/AST/OpenACCClause.cpp | 24 +
clang/lib/AST/StmtProfile.cpp | 6 +
clang/lib/AST/TextNodeDumper.cpp | 1 +
clang/lib/Parse/ParseOpenACC.cpp | 118 ++--
clang/lib/Sema/SemaOpenACC.cpp | 73 +++
clang/lib/Sema/TreeTransform.h | 32 +-
clang/lib/Serialization/ASTReader.cpp | 15 +-
clang/lib/Serialization/ASTWriter.cpp | 13 +-
.../ParserOpenACC/parse-cache-construct.c | 2 +
clang/test/ParserOpenACC/parse-clauses.c | 60 +-
.../compute-construct-private-clause.c | 138 +++++
.../compute-construct-private-clause.cpp | 161 +++++
.../compute-construct-varlist-ast.cpp | 552 ++++++++++++++++++
clang/tools/libclang/CIndex.cpp | 9 +
21 files changed, 1226 insertions(+), 118 deletions(-)
create mode 100644 clang/test/SemaOpenACC/compute-construct-private-clause.c
create mode 100644 clang/test/SemaOpenACC/compute-construct-private-clause.cpp
create mode 100644 clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h
index 277a351c49fcb8..dafcad4179a37e 100644
--- a/clang/include/clang/AST/OpenACCClause.h
+++ b/clang/include/clang/AST/OpenACCClause.h
@@ -156,51 +156,50 @@ class OpenACCSelfClause : public OpenACCClauseWithCondition {
Expr *ConditionExpr, SourceLocation EndLoc);
};
-/// Represents a clause that has one or more IntExprs. It does not own the
-/// IntExprs, but provides 'children' and other accessors.
-class OpenACCClauseWithIntExprs : public OpenACCClauseWithParams {
- MutableArrayRef<Expr *> IntExprs;
+/// Represents a clause that has one or more expressions associated with it.
+class OpenACCClauseWithExprs : public OpenACCClauseWithParams {
+ MutableArrayRef<Expr *> Exprs;
protected:
- OpenACCClauseWithIntExprs(OpenACCClauseKind K, SourceLocation BeginLoc,
- SourceLocation LParenLoc, SourceLocation EndLoc)
+ OpenACCClauseWithExprs(OpenACCClauseKind K, SourceLocation BeginLoc,
+ SourceLocation LParenLoc, SourceLocation EndLoc)
: OpenACCClauseWithParams(K, BeginLoc, LParenLoc, EndLoc) {}
/// Used only for initialization, the leaf class can initialize this to
/// trailing storage.
- void setIntExprs(MutableArrayRef<Expr *> NewIntExprs) {
- assert(IntExprs.empty() && "Cannot change IntExprs list");
- IntExprs = NewIntExprs;
+ void setExprs(MutableArrayRef<Expr *> NewExprs) {
+ assert(Exprs.empty() && "Cannot change Exprs list");
+ Exprs = NewExprs;
}
- /// Gets the entire list of integer expressions, but leave it to the
+ /// Gets the entire list of expressions, but leave it to the
/// individual clauses to expose this how they'd like.
- llvm::ArrayRef<Expr *> getIntExprs() const { return IntExprs; }
+ llvm::ArrayRef<Expr *> getExprs() const { return Exprs; }
public:
child_range children() {
- return child_range(reinterpret_cast<Stmt **>(IntExprs.begin()),
- reinterpret_cast<Stmt **>(IntExprs.end()));
+ return child_range(reinterpret_cast<Stmt **>(Exprs.begin()),
+ reinterpret_cast<Stmt **>(Exprs.end()));
}
const_child_range children() const {
child_range Children =
- const_cast<OpenACCClauseWithIntExprs *>(this)->children();
+ const_cast<OpenACCClauseWithExprs *>(this)->children();
return const_child_range(Children.begin(), Children.end());
}
};
class OpenACCNumGangsClause final
- : public OpenACCClauseWithIntExprs,
+ : public OpenACCClauseWithExprs,
public llvm::TrailingObjects<OpenACCNumGangsClause, Expr *> {
OpenACCNumGangsClause(SourceLocation BeginLoc, SourceLocation LParenLoc,
ArrayRef<Expr *> IntExprs, SourceLocation EndLoc)
- : OpenACCClauseWithIntExprs(OpenACCClauseKind::NumGangs, BeginLoc,
- LParenLoc, EndLoc) {
+ : OpenACCClauseWithExprs(OpenACCClauseKind::NumGangs, BeginLoc, LParenLoc,
+ EndLoc) {
std::uninitialized_copy(IntExprs.begin(), IntExprs.end(),
getTrailingObjects<Expr *>());
- setIntExprs(MutableArrayRef(getTrailingObjects<Expr *>(), IntExprs.size()));
+ setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), IntExprs.size()));
}
public:
@@ -209,35 +208,35 @@ class OpenACCNumGangsClause final
ArrayRef<Expr *> IntExprs, SourceLocation EndLoc);
llvm::ArrayRef<Expr *> getIntExprs() {
- return OpenACCClauseWithIntExprs::getIntExprs();
+ return OpenACCClauseWithExprs::getExprs();
}
llvm::ArrayRef<Expr *> getIntExprs() const {
- return OpenACCClauseWithIntExprs::getIntExprs();
+ return OpenACCClauseWithExprs::getExprs();
}
};
/// Represents one of a handful of clauses that have a single integer
/// expression.
-class OpenACCClauseWithSingleIntExpr : public OpenACCClauseWithIntExprs {
+class OpenACCClauseWithSingleIntExpr : public OpenACCClauseWithExprs {
Expr *IntExpr;
protected:
OpenACCClauseWithSingleIntExpr(OpenACCClauseKind K, SourceLocation BeginLoc,
SourceLocation LParenLoc, Expr *IntExpr,
SourceLocation EndLoc)
- : OpenACCClauseWithIntExprs(K, BeginLoc, LParenLoc, EndLoc),
+ : OpenACCClauseWithExprs(K, BeginLoc, LParenLoc, EndLoc),
IntExpr(IntExpr) {
- setIntExprs(MutableArrayRef<Expr *>{&this->IntExpr, 1});
+ setExprs(MutableArrayRef<Expr *>{&this->IntExpr, 1});
}
public:
- bool hasIntExpr() const { return !getIntExprs().empty(); }
+ bool hasIntExpr() const { return !getExprs().empty(); }
const Expr *getIntExpr() const {
- return hasIntExpr() ? getIntExprs()[0] : nullptr;
+ return hasIntExpr() ? getExprs()[0] : nullptr;
}
- Expr *getIntExpr() { return hasIntExpr() ? getIntExprs()[0] : nullptr; };
+ Expr *getIntExpr() { return hasIntExpr() ? getExprs()[0] : nullptr; };
};
class OpenACCNumWorkersClause : public OpenACCClauseWithSingleIntExpr {
@@ -261,6 +260,40 @@ class OpenACCVectorLengthClause : public OpenACCClauseWithSingleIntExpr {
Expr *IntExpr, SourceLocation EndLoc);
};
+/// Represents a clause with one or more 'var' objects, represented as an expr,
+/// as its arguments. Var-list is expected to be stored in trailing storage.
+/// For now, we're just storing the original expression in its entirety, unlike
+/// OMP which has to do a bunch of work to create a private.
+class OpenACCClauseWithVarList : public OpenACCClauseWithExprs {
+protected:
+ OpenACCClauseWithVarList(OpenACCClauseKind K, SourceLocation BeginLoc,
+ SourceLocation LParenLoc, SourceLocation EndLoc)
+ : OpenACCClauseWithExprs(K, BeginLoc, LParenLoc, EndLoc) {}
+
+public:
+ ArrayRef<Expr *> getVarList() { return getExprs(); }
+ ArrayRef<Expr *> getVarList() const { return getExprs(); }
+};
+
+class OpenACCPrivateClause final
+ : public OpenACCClauseWithVarList,
+ public llvm::TrailingObjects<OpenACCPrivateClause, Expr *> {
+
+ OpenACCPrivateClause(SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc)
+ : OpenACCClauseWithVarList(OpenACCClauseKind::Private, BeginLoc,
+ LParenLoc, EndLoc) {
+ std::uninitialized_copy(VarList.begin(), VarList.end(),
+ getTrailingObjects<Expr *>());
+ setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size()));
+ }
+
+public:
+ static OpenACCPrivateClause *
+ Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc);
+};
+
template <class Impl> class OpenACCClauseVisitor {
Impl &getDerived() { return static_cast<Impl &>(*this); }
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index f72d5c252b863e..b58ef1d451a3ae 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12305,4 +12305,7 @@ def err_acc_num_gangs_num_args
"OpenACC 'num_gangs' "
"%select{|clause: '%1' directive expects maximum of %2, %3 were "
"provided}0">;
+def err_acc_not_a_var_ref
+ : Error<"OpenACC variable is not a valid variable name, sub-array, array "
+ "element, or composite variable member">;
} // end of sema component.
diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def
index dd5792e7ca8c39..6c3c2db66ef0cf 100644
--- a/clang/include/clang/Basic/OpenACCClauses.def
+++ b/clang/include/clang/Basic/OpenACCClauses.def
@@ -20,6 +20,7 @@ VISIT_CLAUSE(If)
VISIT_CLAUSE(Self)
VISIT_CLAUSE(NumGangs)
VISIT_CLAUSE(NumWorkers)
+VISIT_CLAUSE(Private)
VISIT_CLAUSE(VectorLength)
#undef VISIT_CLAUSE
diff --git a/clang/include/clang/Parse/Parser.h b/clang/include/clang/Parse/Parser.h
index fb117bf04087ee..ae82ebd0586f72 100644
--- a/clang/include/clang/Parse/Parser.h
+++ b/clang/include/clang/Parse/Parser.h
@@ -3645,11 +3645,12 @@ class Parser : public CodeCompletionHandler {
ExprResult ParseOpenACCIDExpression();
/// Parses the variable list for the `cache` construct.
void ParseOpenACCCacheVarList();
+
+ using OpenACCVarParseResult = std::pair<ExprResult, OpenACCParseCanContinue>;
/// Parses a single variable in a variable list for OpenACC.
- bool ParseOpenACCVar();
- /// Parses the variable list for the variety of clauses that take a var-list,
- /// including the optional Special Token listed for some,based on clause type.
- bool ParseOpenACCClauseVarList(OpenACCClauseKind Kind);
+ OpenACCVarParseResult ParseOpenACCVar();
+ /// Parses the variable list for the variety of places that take a var-list.
+ llvm::SmallVector<Expr *> ParseOpenACCVarList();
/// Parses any parameters for an OpenACC Clause, including required/optional
/// parens.
OpenACCClauseParseResult
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index da19503c2902fd..9c4ca40e34c5c4 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -48,8 +48,12 @@ class SemaOpenACC : public SemaBase {
SmallVector<Expr *> IntExprs;
};
+ struct VarListDetails {
+ SmallVector<Expr *> VarList;
+ };
+
std::variant<std::monostate, DefaultDetails, ConditionDetails,
- IntExprDetails>
+ IntExprDetails, VarListDetails>
Details = std::monostate{};
public:
@@ -112,6 +116,18 @@ class SemaOpenACC : public SemaBase {
return const_cast<OpenACCParsedClause *>(this)->getIntExprs();
}
+ // Non-const version that permits modifying of the VarList for the purposes
+ // of Sema enforcement.
+ SmallVector<Expr *> &getVarList() {
+ assert(ClauseKind == OpenACCClauseKind::Private &&
+ "Parsed clause kind does not have a var-list");
+ return std::get<VarListDetails>(Details).VarList;
+ }
+
+ ArrayRef<Expr *> getVarList() const {
+ return const_cast<OpenACCParsedClause *>(this)->getVarList();
+ }
+
void setLParenLoc(SourceLocation EndLoc) { LParenLoc = EndLoc; }
void setEndLoc(SourceLocation EndLoc) { ClauseRange.setEnd(EndLoc); }
@@ -147,7 +163,19 @@ class SemaOpenACC : public SemaBase {
ClauseKind == OpenACCClauseKind::NumWorkers ||
ClauseKind == OpenACCClauseKind::VectorLength) &&
"Parsed clause kind does not have a int exprs");
- Details = IntExprDetails{IntExprs};
+ Details = IntExprDetails{std::move(IntExprs)};
+ }
+
+ void setVarListDetails(ArrayRef<Expr *> VarList) {
+ assert(ClauseKind == OpenACCClauseKind::Private &&
+ "Parsed clause kind does not have a var-list");
+ Details = VarListDetails{{VarList.begin(), VarList.end()}};
+ }
+
+ void setVarListDetails(llvm::SmallVector<Expr *> &&VarList) {
+ assert(ClauseKind == OpenACCClauseKind::Private &&
+ "Parsed clause kind does not have a var-list");
+ Details = VarListDetails{std::move(VarList)};
}
};
@@ -194,6 +222,10 @@ class SemaOpenACC : public SemaBase {
ExprResult ActOnIntExpr(OpenACCDirectiveKind DK, OpenACCClauseKind CK,
SourceLocation Loc, Expr *IntExpr);
+ /// Called when encountering a 'var' for OpenACC, ensures it is actually a
+ /// declaration reference to a variable of the correct type.
+ ExprResult ActOnVar(Expr *VarExpr);
+
/// Checks and creates an Array Section used in an OpenACC construct/clause.
ExprResult ActOnArraySectionExpr(Expr *Base, SourceLocation LBLoc,
Expr *LowerBound,
diff --git a/clang/include/clang/Serialization/ASTRecordReader.h b/clang/include/clang/Serialization/ASTRecordReader.h
index 06b80f266a9441..1e11d2d5e42f95 100644
--- a/clang/include/clang/Serialization/ASTRecordReader.h
+++ b/clang/include/clang/Serialization/ASTRecordReader.h
@@ -269,6 +269,9 @@ class ASTRecordReader
/// Read an OpenMP children, advancing Idx.
void readOMPChildren(OMPChildren *Data);
+ /// Read a list of Exprs used for a var-list.
+ llvm::SmallVector<Expr *> readOpenACCVarList();
+
/// Read an OpenACC clause, advancing Idx.
OpenACCClause *readOpenACCClause();
diff --git a/clang/include/clang/Serialization/ASTRecordWriter.h b/clang/include/clang/Serialization/ASTRecordWriter.h
index 1feb8fcbacf772..8b1da49bd4c576 100644
--- a/clang/include/clang/Serialization/ASTRecordWriter.h
+++ b/clang/include/clang/Serialization/ASTRecordWriter.h
@@ -15,6 +15,7 @@
#define LLVM_CLANG_SERIALIZATION_ASTRECORDWRITER_H
#include "clang/AST/AbstractBasicWriter.h"
+#include "clang/AST/OpenACCClause.h"
#include "clang/AST/OpenMPClause.h"
#include "clang/Serialization/ASTWriter.h"
#include "clang/Serialization/SourceLocationEncoding.h"
@@ -293,6 +294,8 @@ class ASTRecordWriter
/// Writes data related to the OpenMP directives.
void writeOMPChildren(OMPChildren *Data);
+ void writeOpenACCVarList(const OpenACCClauseWithVarList *C);
+
/// Writes out a single OpenACC Clause.
void writeOpenACCClause(const OpenACCClause *C);
diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp
index 6cd5b28802187d..208a51c82399cc 100644
--- a/clang/lib/AST/OpenACCClause.cpp
+++ b/clang/lib/AST/OpenACCClause.cpp
@@ -134,6 +134,24 @@ OpenACCNumGangsClause *OpenACCNumGangsClause::Create(const ASTContext &C,
return new (Mem) OpenACCNumGangsClause(BeginLoc, LParenLoc, IntExprs, EndLoc);
}
+OpenACCPrivateClause *OpenACCPrivateClause::Create(const ASTContext &C,
+ SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList,
+ SourceLocation EndLoc) {
+ void *Mem = C.Allocate(
+ OpenACCPrivateClause::totalSizeToAlloc<Expr *>(VarList.size()));
+ return new (Mem) OpenACCPrivateClause(BeginLoc, LParenLoc, VarList, EndLoc);
+}
+
+// ValueDecl *getDeclFromExpr(Expr *RefExpr) {
+// //RefExpr = RefExpr->IgnoreParenImpCasts();
+//
+// ////while (isa<ArraySubscriptExpr, ArraySectionExpr>(RefExpr)) {
+// ////}
+// // TODO:
+// }
+
//===----------------------------------------------------------------------===//
// OpenACC clauses printing methods
//===----------------------------------------------------------------------===//
@@ -166,3 +184,9 @@ void OpenACCClausePrinter::VisitVectorLengthClause(
const OpenACCVectorLengthClause &C) {
OS << "vector_length(" << C.getIntExpr() << ")";
}
+
+void OpenACCClausePrinter::VisitPrivateClause(const OpenACCPrivateClause &C) {
+ OS << "private(";
+ llvm::interleaveComma(C.getVarList(), OS);
+ OS << ")";
+}
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index a95f5c6103e24d..973f6f97bae0bf 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -2509,6 +2509,12 @@ void OpenACCClauseProfiler::VisitNumWorkersClause(
Profiler.VisitStmt(Clause.getIntExpr());
}
+void OpenACCClauseProfiler::VisitPrivateClause(
+ const OpenACCPrivateClause &Clause) {
+ for (auto *E : Clause.getVarList())
+ Profiler.VisitStmt(E);
+}
+
void OpenACCClauseProfiler::VisitVectorLengthClause(
const OpenACCVectorLengthClause &Clause) {
assert(Clause.hasIntExpr() &&
diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp
index 8f0a9a9b0ed0bc..89f50d6dacfd23 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -401,6 +401,7 @@ void TextNodeDumper::Visit(const OpenACCClause *C) {
case OpenACCClauseKind::Self:
case OpenACCClauseKind::NumGangs:
case OpenACCClauseKind::NumWorkers:
+ case OpenACCClauseKind::Private:
case OpenACCClauseKind::VectorLength:
// The condition expression will be printed as a part of the 'children',
// but print 'clause' here so it is clear what is happening from the dump.
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index 29326f5d993a9d..a12ffad699755f 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -86,6 +86,10 @@ OpenACCClauseKind getOpenACCClauseKind(Token Tok) {
if (Tok.is(tok::kw_if))
return OpenACCClauseKind::If;
+ // 'private' is also a keyword, make sure we pare it correctly.
+ if (Tok.is(tok::kw_private))
+ return OpenACCClauseKind::Private;
+
if (!Tok.is(tok::identifier))
return OpenACCClauseKind::Invalid;
@@ -682,28 +686,6 @@ bool Parser::ParseOpenACCIntExprList(OpenACCDirectiveKind DK,
return false;
}
-bool Parser::ParseOpenACCClauseVarList(OpenACCClauseKind Kind) {
- // FIXME: Future clauses will require 'special word' parsing, check for one,
- // then parse it based on whether it is a clause that requires a 'special
- // word'.
- (void)Kind;
-
- // If the var parsing fails, skip until the end of the directive as this is
- // an expression and gets messy if we try to continue otherwise.
- if (ParseOpenACCVar())
- return true;
-
- while (!getCurToken().isOneOf(tok::r_paren, tok::annot_pragma_openacc_end)) {
- ExpectAndConsume(tok::comma);
-
- // If the var parsing fails, skip until the end of the directive as this is
- // an expression and gets messy if we try to continue otherwise.
- if (ParseOpenACCVar())
- return true;
- }
- return false;
-}
-
/// OpenACC 3.3 Section 2.4:
/// The argument to the device_type clause is a comma-separated list of one or
/// more device architecture name identifiers, or an asterisk.
@@ -917,28 +899,19 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
case OpenACCClauseKind::CopyIn:
tryParseAndConsumeSpecialTokenKind(
*this, OpenACCSpecialTokenKind::ReadOnly, ClauseKind);
- if (ParseOpenACCClauseVarList(ClauseKind)) {
- Parens.skipToEnd();
- return OpenACCCanContinue();
- }
+ ParseOpenACCVarList();
break;
case OpenACCClauseKind::Create:
case OpenACCClauseKind::CopyOut:
tryParseAndConsumeSpecialTokenKind(*this, OpenACCSpecialTokenKind::Zero,
ClauseKind);
- if (ParseOpenACCClauseVarList(ClauseKind)) {
- Parens.skipToEnd();
- return OpenACCCanContinue();
- }
+ ParseOpenACCVarList();
break;
case OpenACCClauseKind::Reduction:
// If we're missing a clause-kind (or it is invalid), see if we can parse
// the var-list anyway.
ParseReductionOperator(*this);
- if (ParseOpenACCClauseVarList(ClauseKind)) {
- Parens.skipToEnd();
- return OpenACCCanContinue();
- }
+ ParseOpenACCVarList();
break;
case OpenACCClauseKind::Self:
// The 'self' clause is a var-list instead of a 'condition' in the case of
@@ -958,13 +931,14 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
case OpenACCClauseKind::Link:
case OpenACCClauseKind::NoCreate:
case OpenACCClauseKind::Present:
- case OpenACCClauseKind::Private:
case OpenACCClauseKind::UseDevice:
- if (ParseOpenACCClauseVarList(ClauseKind)) {
- Parens.skipToEnd();
- return OpenACCCanContinue();
- }
+ ParseOpenACCVarList();
+ break;
+ case OpenACCClauseKind::Private: {
+ llvm::SmallVector<Expr *> Vars = ParseOpenACCVarList();
+ ParsedClause.setVarListDetails(std::move(Vars));
break;
+ }
case OpenACCClauseKind::Collapse: {
tryParseAndConsumeSpecialTokenKind(*this, OpenACCSpecialTokenKind::Force,
ClauseKind);
@@ -1227,16 +1201,51 @@ ExprResult Parser::ParseOpenACCBindClauseArgument() {
/// OpenACC 3.3, section 1.6:
/// In this spec, a 'var' (in italics) is one of the following:
-/// - a variable name (a scalar, array, or compisite variable name)
+/// - a variable name (a scalar, array, or composite variable name)
/// - a subarray specification with subscript ranges
/// - an array element
/// - a member of a composite variable
/// - a common block name between slashes (fortran only)
-bool Parser::ParseOpenACCVar() {
+Parser::OpenACCVarParseResult Parser::ParseOpenACCVar() {
OpenACCArraySectionRAII ArraySections(*this);
- ExprResult Res =
- getActions().CorrectDelayedTyposInExpr(ParseAssignmentExpression());
- return Res.isInvalid();
+
+ ExprResult Res = ParseAssignmentExpression();
+ if (!Res.isUsable())
+ return {Res, OpenACCParseCanContinue::Cannot};
+
+ Res = getActions().CorrectDelayedTyposInExpr(Res.get());
+ if (!Res.isUsable())
+ return {Res, OpenACCParseCanContinue::Can};
+
+ Res = getActions().OpenACC().ActOnVar(Res.get());
+
+ return {Res, OpenACCParseCanContinue::Can};
+}
+
+llvm::SmallVector<Expr *> Parser::ParseOpenACCVarList() {
+ llvm::SmallVector<Expr *> Vars;
+
+ auto [Res, CanContinue] = ParseOpenACCVar();
+ if (Res.isUsable()) {
+ Vars.push_back(Res.get());
+ } else if (CanContinue == OpenACCParseCanContinue::Cannot) {
+ SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end, StopBeforeMatch);
+ return Vars;
+ }
+
+ while (!getCurToken().isOneOf(tok::r_paren, tok::annot_pragma_openacc_end)) {
+ ExpectAndConsume(tok::comma);
+
+ auto [Res, CanContinue] = ParseOpenACCVar();
+
+ if (Res.isUsable()) {
+ Vars.push_back(Res.get());
+ } else if (CanContinue == OpenACCParseCanContinue::Cannot) {
+ SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end, StopBeforeMatch);
+ return Vars;
+ }
+ }
+ return Vars;
}
/// OpenACC 3.3, section 2.10:
@@ -1259,24 +1268,9 @@ void Parser::ParseOpenACCCacheVarList() {
// Sema/AST generation.
}
- bool FirstArray = true;
- while (!getCurToken().isOneOf(tok::r_paren, tok::annot_pragma_openacc_end)) {
- if (!FirstArray)
- ExpectAndConsume(tok::comma);
- FirstArray = false;
-
- // OpenACC 3.3, section 2.10:
- // A 'var' in a cache directive must be a single array element or a simple
- // subarray. In C and C++, a simple subarray is an array name followed by
- // an extended array range specification in brackets, with a start and
- // length such as:
- //
- // arr[lower:length]
- //
- if (ParseOpenACCVar())
- SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end, tok::comma,
- StopBeforeMatch);
- }
+ // ParseOpenACCVarList should leave us before a r-paren, so no need to skip
+ // anything here.
+ ParseOpenACCVarList();
}
Parser::OpenACCDirectiveParseInfo Parser::ParseOpenACCDirective() {
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index d5cfe82a5d7098..3ea81e0497c203 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -103,6 +103,18 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind,
default:
return false;
}
+ case OpenACCClauseKind::Private:
+ switch (DirectiveKind) {
+ case OpenACCDirectiveKind::Parallel:
+ case OpenACCDirectiveKind::Serial:
+ case OpenACCDirectiveKind::Loop:
+ case OpenACCDirectiveKind::ParallelLoop:
+ case OpenACCDirectiveKind::SerialLoop:
+ case OpenACCDirectiveKind::KernelsLoop:
+ return true;
+ default:
+ return false;
+ }
default:
// Do nothing so we can go to the 'unimplemented' diagnostic instead.
return true;
@@ -303,6 +315,21 @@ SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses,
getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(),
Clause.getIntExprs()[0], Clause.getEndLoc());
}
+ case OpenACCClauseKind::Private: {
+ // Restrictions only properly implemented on 'compute' constructs, and
+ // 'compute' constructs are the only construct that can do anything with
+ // this yet, so skip/treat as unimplemented in this case.
+ if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind()))
+ break;
+
+ // ActOnVar ensured that everything is a valid variable reference, so there
+ // really isn't anything to do here. GCC does some duplicate-finding, though
+ // it isn't apparent in the standard where this is justified.
+
+ return OpenACCPrivateClause::Create(
+ getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(),
+ Clause.getVarList(), Clause.getEndLoc());
+ }
default:
break;
}
@@ -423,6 +450,52 @@ ExprResult SemaOpenACC::ActOnIntExpr(OpenACCDirectiveKind DK,
return IntExpr;
}
+ExprResult SemaOpenACC::ActOnVar(Expr *VarExpr) {
+ // We still need to retain the array subscript/subarray exprs, so work on a
+ // copy.
+ Expr *CurVarExpr = VarExpr->IgnoreParenImpCasts();
+
+ // Sub-arrays/subscript-exprs are fine as long as the base is a
+ // VarExpr/MemberExpr. So strip all of those off.
+ while (isa<ArraySectionExpr, ArraySubscriptExpr>(CurVarExpr)) {
+ if (auto *SubScrpt = dyn_cast<ArraySubscriptExpr>(CurVarExpr))
+ CurVarExpr = SubScrpt->getBase()->IgnoreParenImpCasts();
+ else
+ CurVarExpr =
+ cast<ArraySectionExpr>(CurVarExpr)->getBase()->IgnoreParenImpCasts();
+ }
+
+ // References to a VarDecl are fine.
+ if (const auto *DRE = dyn_cast<DeclRefExpr>(CurVarExpr)) {
+ if (isa<VarDecl, NonTypeTemplateParmDecl>(
+ DRE->getDecl()->getCanonicalDecl()))
+ return VarExpr;
+ }
+
+ // A MemberExpr that references a Field is valid.
+ if (const auto *ME = dyn_cast<MemberExpr>(CurVarExpr)) {
+ if (isa<FieldDecl>(ME->getMemberDecl()->getCanonicalDecl()))
+ return VarExpr;
+ }
+
+ // Referring to 'this' is always OK.
+ if (isa<CXXThisExpr>(CurVarExpr))
+ return VarExpr;
+
+ // Nothing really we can do here, as these are dependent. So just return they
+ // are valid.
+ if (isa<DependentScopeDeclRefExpr, CXXDependentScopeMemberExpr>(CurVarExpr))
+ return VarExpr;
+
+ // There isn't really anything we can do in the case of a recovery expr, so
+ // skip the diagnostic rather than produce a confusing diagnostic.
+ if (isa<RecoveryExpr>(CurVarExpr))
+ return ExprError();
+
+ Diag(VarExpr->getExprLoc(), diag::err_acc_not_a_var_ref);
+ return ExprError();
+}
+
ExprResult SemaOpenACC::ActOnArraySectionExpr(Expr *Base, SourceLocation LBLoc,
Expr *LowerBound,
SourceLocation ColonLoc,
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index f47bc219e6fa32..1337596621f64f 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -11104,13 +11104,15 @@ template <typename Derived>
class OpenACCClauseTransform final
: public OpenACCClauseVisitor<OpenACCClauseTransform<Derived>> {
TreeTransform<Derived> &Self;
+ ArrayRef<const OpenACCClause *> ExistingClauses;
SemaOpenACC::OpenACCParsedClause &ParsedClause;
OpenACCClause *NewClause = nullptr;
public:
OpenACCClauseTransform(TreeTransform<Derived> &Self,
+ ArrayRef<const OpenACCClause *> ExistingClauses,
SemaOpenACC::OpenACCParsedClause &PC)
- : Self(Self), ParsedClause(PC) {}
+ : Self(Self), ExistingClauses(ExistingClauses), ParsedClause(PC) {}
OpenACCClause *CreatedClause() const { return NewClause; }
@@ -11196,6 +11198,31 @@ void OpenACCClauseTransform<Derived>::VisitNumGangsClause(
ParsedClause.getLParenLoc(), ParsedClause.getIntExprs(),
ParsedClause.getEndLoc());
}
+
+template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitPrivateClause(
+ const OpenACCPrivateClause &C) {
+ llvm::SmallVector<Expr *> InstantiatedVarList;
+
+ for (Expr *CurVar : C.getVarList()) {
+ ExprResult Res = Self.TransformExpr(CurVar);
+
+ if (!Res.isUsable())
+ return;
+
+ Res = Self.getSema().OpenACC().ActOnVar(Res.get());
+
+ if (Res.isUsable())
+ InstantiatedVarList.push_back(Res.get());
+ }
+ ParsedClause.setVarListDetails(std::move(InstantiatedVarList));
+
+ NewClause = OpenACCPrivateClause::Create(
+ Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
+ ParsedClause.getLParenLoc(), ParsedClause.getVarList(),
+ ParsedClause.getEndLoc());
+}
+
template <typename Derived>
void OpenACCClauseTransform<Derived>::VisitNumWorkersClause(
const OpenACCNumWorkersClause &C) {
@@ -11254,7 +11281,8 @@ OpenACCClause *TreeTransform<Derived>::TransformOpenACCClause(
if (const auto *WithParms = dyn_cast<OpenACCClauseWithParams>(OldClause))
ParsedClause.setLParenLoc(WithParms->getLParenLoc());
- OpenACCClauseTransform<Derived> Transform{*this, ParsedClause};
+ OpenACCClauseTransform<Derived> Transform{*this, ExistingClauses,
+ ParsedClause};
Transform.Visit(OldClause);
return Transform.CreatedClause();
diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp
index 0ef57a3ea804ef..143fbc7feb3ab7 100644
--- a/clang/lib/Serialization/ASTReader.cpp
+++ b/clang/lib/Serialization/ASTReader.cpp
@@ -11768,6 +11768,14 @@ void ASTRecordReader::readOMPChildren(OMPChildren *Data) {
Data->getChildren()[I] = readStmt();
}
+SmallVector<Expr *> ASTRecordReader::readOpenACCVarList() {
+ unsigned NumVars = readInt();
+ llvm::SmallVector<Expr *> VarList;
+ for (unsigned I = 0; I < NumVars; ++I)
+ VarList.push_back(readSubExpr());
+ return VarList;
+}
+
OpenACCClause *ASTRecordReader::readOpenACCClause() {
OpenACCClauseKind ClauseKind = readEnum<OpenACCClauseKind>();
SourceLocation BeginLoc = readSourceLocation();
@@ -11813,6 +11821,12 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
return OpenACCVectorLengthClause::Create(getContext(), BeginLoc, LParenLoc,
IntExpr, EndLoc);
}
+ case OpenACCClauseKind::Private: {
+ SourceLocation LParenLoc = readSourceLocation();
+ llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
+ return OpenACCPrivateClause::Create(getContext(), BeginLoc, LParenLoc,
+ VarList, EndLoc);
+ }
case OpenACCClauseKind::Finalize:
case OpenACCClauseKind::IfPresent:
case OpenACCClauseKind::Seq:
@@ -11834,7 +11848,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
case OpenACCClauseKind::Link:
case OpenACCClauseKind::NoCreate:
case OpenACCClauseKind::Present:
- case OpenACCClauseKind::Private:
case OpenACCClauseKind::CopyOut:
case OpenACCClauseKind::CopyIn:
case OpenACCClauseKind::Create:
diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp
index 0408eeb6a95b00..c886fc59587879 100644
--- a/clang/lib/Serialization/ASTWriter.cpp
+++ b/clang/lib/Serialization/ASTWriter.cpp
@@ -7645,6 +7645,12 @@ void ASTRecordWriter::writeOMPChildren(OMPChildren *Data) {
AddStmt(Data->getChildren()[I]);
}
+void ASTRecordWriter::writeOpenACCVarList(const OpenACCClauseWithVarList *C) {
+ writeUInt32(C->getVarList().size());
+ for (Expr *E : C->getVarList())
+ AddStmt(E);
+}
+
void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
writeEnum(C->getClauseKind());
writeSourceLocation(C->getBeginLoc());
@@ -7691,6 +7697,12 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
AddStmt(const_cast<Expr *>(NWC->getIntExpr()));
return;
}
+ case OpenACCClauseKind::Private: {
+ const auto *PC = cast<OpenACCPrivateClause>(C);
+ writeSourceLocation(PC->getLParenLoc());
+ writeOpenACCVarList(PC);
+ return;
+ }
case OpenACCClauseKind::Finalize:
case OpenACCClauseKind::IfPresent:
case OpenACCClauseKind::Seq:
@@ -7712,7 +7724,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
case OpenACCClauseKind::Link:
case OpenACCClauseKind::NoCreate:
case OpenACCClauseKind::Present:
- case OpenACCClauseKind::Private:
case OpenACCClauseKind::CopyOut:
case OpenACCClauseKind::CopyIn:
case OpenACCClauseKind::Create:
diff --git a/clang/test/ParserOpenACC/parse-cache-construct.c b/clang/test/ParserOpenACC/parse-cache-construct.c
index fd161c03c09f75..de26fc2b277a6b 100644
--- a/clang/test/ParserOpenACC/parse-cache-construct.c
+++ b/clang/test/ParserOpenACC/parse-cache-construct.c
@@ -25,11 +25,13 @@ void func() {
}
for (int i = 0; i < 10; ++i) {
+ // expected-error at +2{{expected expression}}
// expected-warning at +1{{OpenACC construct 'cache' not yet implemented, pragma ignored}}
#pragma acc cache()
}
for (int i = 0; i < 10; ++i) {
+ // expected-error at +3{{expected expression}}
// expected-error at +2{{invalid OpenACC clause 'clause'}}
// expected-warning at +1{{OpenACC construct 'cache' not yet implemented, pragma ignored}}
#pragma acc cache() clause-list
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index ee2cb2d1501dea..8a439a5ccd4bdc 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -405,7 +405,10 @@ void SelfUpdate() {
#pragma acc update self
for(;;){}
- // expected-error at +3{{use of undeclared identifier 'zero'}}
+ // expected-error at +6{{use of undeclared identifier 'zero'}}
+ // expected-error at +5{{expected ','}}
+ // expected-error at +4{{expected expression}}
+ // expected-warning at +3{{OpenACC clause 'self' not yet implemented, clause ignored}}
// expected-warning at +2{{OpenACC clause 'seq' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC construct 'update' not yet implemented, pragma ignored}}
#pragma acc update self(zero : s.array[s.value : 5], s.value), seq
@@ -450,11 +453,13 @@ void VarListClauses() {
#pragma acc serial copy(, seq
for(;;){}
- // expected-error at +1{{expected expression}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'copy' not yet implemented, clause ignored}}
#pragma acc serial copy()
for(;;){}
- // expected-error at +2{{expected expression}}
+ // expected-error at +3{{expected expression}}
+ // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(), seq
for(;;){}
@@ -494,24 +499,28 @@ void VarListClauses() {
#pragma acc serial copy(HasMem.MemArr[1:3].array[1:2]), seq
for(;;){}
- // expected-error at +2{{expected expression}}
+ // expected-error at +3{{expected expression}}
+ // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[:]), seq
for(;;){}
- // expected-error at +2{{expected expression}}
+ // expected-error at +3{{expected expression}}
+ // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[::]), seq
for(;;){}
- // expected-error at +4{{expected expression}}
- // expected-error at +3{{expected ']'}}
- // expected-note at +2{{to match this '['}}
+ // expected-error at +5{{expected expression}}
+ // expected-error at +4{{expected ']'}}
+ // expected-note at +3{{to match this '['}}
+ // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[: :]), seq
for(;;){}
- // expected-error at +2{{expected expression}}
+ // expected-error at +3{{expected expression}}
+ // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[3:]), seq
for(;;){}
@@ -582,13 +591,11 @@ void VarListClauses() {
#pragma acc serial detach(s.array[s.value : 5], s.value), seq
for(;;){}
- // expected-error at +3{{expected ','}}
- // expected-warning at +2{{OpenACC clause 'private' not yet implemented, clause ignored}}
+ // expected-error at +2{{expected ','}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial private(s.array[s.value] s.array[s.value :5] ), seq
for(;;){}
- // expected-warning at +2{{OpenACC clause 'private' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial private(s.array[s.value : 5], s.value), seq
for(;;){}
@@ -691,7 +698,9 @@ void VarListClauses() {
#pragma acc serial copyout(zero : s.array[s.value : 5], s.value), seq
for(;;){}
- // expected-error at +2{{use of undeclared identifier 'zero'}}
+ // expected-error at +4{{use of undeclared identifier 'zero'}}
+ // expected-error at +3{{expected ','}}
+ // expected-warning at +2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copyout(zero s.array[s.value : 5], s.value), seq
for(;;){}
@@ -714,7 +723,9 @@ void VarListClauses() {
#pragma acc serial copyout(invalid:s.array[s.value : 5], s.value), seq
for(;;){}
- // expected-error at +2{{use of undeclared identifier 'invalid'}}
+ // expected-error at +4{{use of undeclared identifier 'invalid'}}
+ // expected-error at +3{{expected ','}}
+ // expected-warning at +2{{OpenACC clause 'copyout' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copyout(invalid s.array[s.value : 5], s.value), seq
for(;;){}
@@ -740,7 +751,9 @@ void VarListClauses() {
#pragma acc serial create(zero : s.array[s.value : 5], s.value), seq
for(;;){}
- // expected-error at +2{{use of undeclared identifier 'zero'}}
+ // expected-error at +4{{use of undeclared identifier 'zero'}}
+ // expected-error at +3{{expected ','}}
+ // expected-warning at +2{{OpenACC clause 'create' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial create(zero s.array[s.value : 5], s.value), seq
for(;;){}
@@ -763,7 +776,9 @@ void VarListClauses() {
#pragma acc serial create(invalid:s.array[s.value : 5], s.value), seq
for(;;){}
- // expected-error at +2{{use of undeclared identifier 'invalid'}}
+ // expected-error at +4{{use of undeclared identifier 'invalid'}}
+ // expected-error at +3{{expected ','}}
+ // expected-warning at +2{{OpenACC clause 'create' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial create(invalid s.array[s.value : 5], s.value), seq
for(;;){}
@@ -789,7 +804,9 @@ void VarListClauses() {
#pragma acc serial copyin(readonly : s.array[s.value : 5], s.value), seq
for(;;){}
- // expected-error at +2{{use of undeclared identifier 'readonly'}}
+ // expected-error at +4{{use of undeclared identifier 'readonly'}}
+ // expected-error at +3{{expected ','}}
+ // expected-warning at +2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copyin(readonly s.array[s.value : 5], s.value), seq
for(;;){}
@@ -812,7 +829,9 @@ void VarListClauses() {
#pragma acc serial copyin(invalid:s.array[s.value : 5], s.value), seq
for(;;){}
- // expected-error at +2{{use of undeclared identifier 'invalid'}}
+ // expected-error at +4{{use of undeclared identifier 'invalid'}}
+ // expected-error at +3{{expected ','}}
+ // expected-warning at +2{{OpenACC clause 'copyin' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copyin(invalid s.array[s.value : 5], s.value), seq
for(;;){}
@@ -823,8 +842,9 @@ void ReductionClauseParsing() {
// expected-error at +1{{expected '('}}
#pragma acc serial reduction
for(;;){}
- // expected-error at +2{{missing reduction operator, expected '+', '*', 'max', 'min', '&', '|', '^', '&&', or '||', follwed by a ':'}}
- // expected-error at +1{{expected expression}}
+ // expected-error at +3{{missing reduction operator, expected '+', '*', 'max', 'min', '&', '|', '^', '&&', or '||', follwed by a ':'}}
+ // expected-error at +2{{expected expression}}
+ // expected-warning at +1{{OpenACC clause 'reduction' not yet implemented, clause ignored}}
#pragma acc serial reduction()
for(;;){}
// expected-error at +2{{missing reduction operator, expected '+', '*', 'max', 'min', '&', '|', '^', '&&', or '||', follwed by a ':'}}
diff --git a/clang/test/SemaOpenACC/compute-construct-private-clause.c b/clang/test/SemaOpenACC/compute-construct-private-clause.c
new file mode 100644
index 00000000000000..959e1175b5e156
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-private-clause.c
@@ -0,0 +1,138 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+struct Incomplete;
+enum SomeE{ A };
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ enum SomeE EnumMember;
+ void *PointerMember;
+} Complete;
+
+int GlobalInt;
+float GlobalArray[5];
+void *GlobalPointer;
+Complete GlobalComposite;
+
+void uses(int IntParam, void *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+ int LocalInt;
+ void *LocalPointer;
+ float LocalArray[5];
+ Complete LocalComposite;
+
+ // Check Appertainment:
+#pragma acc parallel private(LocalInt)
+ while(1);
+#pragma acc serial private(LocalInt)
+ while(1);
+ // expected-error at +1{{OpenACC 'private' clause is not valid on 'kernels' directive}}
+#pragma acc kernels private(LocalInt)
+ while(1);
+
+ // Valid cases:
+#pragma acc parallel private(LocalInt, LocalPointer, LocalArray)
+ while(1);
+#pragma acc parallel private(LocalArray)
+ while(1);
+ // TODO OpenACC: Fix array sections, this should be allowed.
+ // expected-error at +1{{expected expression}}
+#pragma acc parallel private(LocalArray[:])
+ while(1);
+#pragma acc parallel private(LocalArray[:5])
+ while(1);
+ // TODO OpenACC: Fix array sections, this should be allowed.
+ // expected-error at +1{{expected expression}}
+#pragma acc parallel private(LocalArray[2:])
+ while(1);
+#pragma acc parallel private(LocalArray[2:5])
+ while(1);
+#pragma acc parallel private(LocalArray[2])
+ while(1);
+#pragma acc parallel private(LocalComposite)
+ while(1);
+#pragma acc parallel private(LocalComposite.EnumMember)
+ while(1);
+#pragma acc parallel private(LocalComposite.ScalarMember)
+ while(1);
+#pragma acc parallel private(LocalComposite.ArrayMember)
+ while(1);
+#pragma acc parallel private(LocalComposite.ArrayMember[5])
+ while(1);
+#pragma acc parallel private(LocalComposite.PointerMember)
+ while(1);
+#pragma acc parallel private(GlobalInt, GlobalArray, GlobalPointer, GlobalComposite)
+ while(1);
+#pragma acc parallel private(GlobalArray[2], GlobalPointer[2], GlobalComposite.CompositeMember.A)
+ while(1);
+#pragma acc parallel private(LocalComposite, GlobalComposite)
+ while(1);
+#pragma acc parallel private(IntParam, PointerParam, ArrayParam, CompositeParam)
+ while(1);
+#pragma acc parallel private(PointerParam[IntParam], ArrayParam[IntParam], CompositeParam.CompositeMember.A)
+ while(1);
+
+#pragma acc parallel private(LocalArray) private(LocalArray[2])
+ while(1);
+
+#pragma acc parallel private(LocalArray, LocalArray[2])
+ while(1);
+
+#pragma acc parallel private(LocalComposite, LocalComposite.ScalarMember)
+ while(1);
+
+#pragma acc parallel private(LocalComposite.CompositeMember.A, LocalComposite.ScalarMember)
+ while(1);
+
+#pragma acc parallel private(LocalComposite.CompositeMember.A) private(LocalComposite.ScalarMember)
+ while(1);
+
+ Complete LocalComposite2;
+#pragma acc parallel private(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember)
+ while(1);
+
+ // Invalid cases, arbitrary expressions.
+ struct Incomplete *I;
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(*I)
+ while(1);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(GlobalInt + IntParam)
+ while(1);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(+GlobalInt)
+ while(1);
+
+ // TODO OpenACC: Fix array sections, this should be allowed.
+ // expected-error at +1{{expected expression}}
+#pragma acc parallel private(PointerParam[:])
+ while(1);
+#pragma acc parallel private(PointerParam[:5])
+ while(1);
+#pragma acc parallel private(PointerParam[:IntParam])
+ while(1);
+ // TODO OpenACC: Fix array sections, this should be allowed.
+ // expected-error at +1{{expected expression}}
+#pragma acc parallel private(PointerParam[2:])
+ while(1);
+#pragma acc parallel private(PointerParam[2:5])
+ while(1);
+#pragma acc parallel private(PointerParam[2])
+ while(1);
+ // TODO OpenACC: Fix array sections, this should be allowed.
+ // expected-error at +1{{expected expression}}
+#pragma acc parallel private(ArrayParam[:])
+ while(1);
+#pragma acc parallel private(ArrayParam[:5])
+ while(1);
+#pragma acc parallel private(ArrayParam[:IntParam])
+ while(1);
+ // TODO OpenACC: Fix array sections, this should be allowed.
+ // expected-error at +1{{expected expression}}
+#pragma acc parallel private(ArrayParam[2:])
+ while(1);
+#pragma acc parallel private(ArrayParam[2:5])
+ while(1);
+#pragma acc parallel private(ArrayParam[2])
+ while(1);
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-private-clause.cpp b/clang/test/SemaOpenACC/compute-construct-private-clause.cpp
new file mode 100644
index 00000000000000..4dd4e0d8029d67
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-private-clause.cpp
@@ -0,0 +1,161 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+struct Incomplete;
+enum SomeE{};
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ SomeE EnumMember;
+ char *PointerMember;
+} Complete;
+
+int GlobalInt;
+float GlobalArray[5];
+char *GlobalPointer;
+Complete GlobalComposite;
+
+void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) {
+ int LocalInt;
+ char *LocalPointer;
+ float LocalArray[5];
+ Complete LocalComposite;
+
+ // Check Appertainment:
+
+#pragma acc parallel private(LocalInt)
+ while(true);
+#pragma acc serial private(LocalInt)
+ while(true);
+ // expected-error at +1{{OpenACC 'private' clause is not valid on 'kernels' directive}}
+#pragma acc kernels private(LocalInt)
+ while(true);
+
+ // Valid cases:
+#pragma acc parallel private(LocalInt, LocalPointer, LocalArray)
+ while(true);
+#pragma acc parallel private(LocalArray)
+ while(true);
+#pragma acc parallel private(LocalArray[2])
+ while(true);
+#pragma acc parallel private(LocalComposite)
+ while(true);
+#pragma acc parallel private(LocalComposite.EnumMember)
+ while(true);
+#pragma acc parallel private(LocalComposite.ScalarMember)
+ while(true);
+#pragma acc parallel private(LocalComposite.ArrayMember)
+ while(true);
+#pragma acc parallel private(LocalComposite.ArrayMember[5])
+ while(true);
+#pragma acc parallel private(LocalComposite.PointerMember)
+ while(true);
+#pragma acc parallel private(GlobalInt, GlobalArray, GlobalPointer, GlobalComposite)
+ while(true);
+#pragma acc parallel private(GlobalArray[2], GlobalPointer[2], GlobalComposite.CompositeMember.A)
+ while(true);
+#pragma acc parallel private(LocalComposite, GlobalComposite)
+ while(true);
+#pragma acc parallel private(IntParam, PointerParam, ArrayParam, CompositeParam) private(IntParamRef)
+ while(true);
+#pragma acc parallel private(PointerParam[IntParam], ArrayParam[IntParam], CompositeParam.CompositeMember.A)
+ while(true);
+
+
+ // Invalid cases, arbitrary expressions.
+ Incomplete *I;
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(*I)
+ while(true);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(GlobalInt + IntParam)
+ while(true);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(+GlobalInt)
+ while(true);
+}
+
+template<typename T, unsigned I, typename V>
+void TemplUses(T t, T (&arrayT)[I], V TemplComp) {
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(+t)
+ while(true);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(+I)
+ while(true);
+
+ // NTTP's are only valid if it is a reference to something.
+ // expected-error at +2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+ // expected-note@#TEMPL_USES_INST{{in instantiation of}}
+#pragma acc parallel private(I)
+ while(true);
+
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private(t, I)
+ while(true);
+
+#pragma acc parallel private(arrayT)
+ while(true);
+
+#pragma acc parallel private(TemplComp)
+ while(true);
+
+#pragma acc parallel private(TemplComp.PointerMember[5])
+ while(true);
+
+#pragma acc parallel private(TemplComp.PointerMember[5]) private(TemplComp)
+ while(true);
+
+ int *Pointer;
+#pragma acc parallel private(Pointer[:I])
+ while(true);
+#pragma acc parallel private(Pointer[:t])
+ while(true);
+ // TODO OpenACC: When fixing sub-arrays, this should be permitted}}
+ // expected-error at +1{{expected expression}}
+#pragma acc parallel private(Pointer[1:])
+ while(true);
+}
+
+template<unsigned I, auto &NTTP_REF>
+void NTTP() {
+ // NTTP's are only valid if it is a reference to something.
+ // expected-error at +2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+ // expected-note@#NTTP_INST{{in instantiation of}}
+#pragma acc parallel private(I)
+ while(true);
+
+#pragma acc parallel private(NTTP_REF)
+ while(true);
+}
+
+struct S {
+ int ThisMember;
+ int ThisMemberArray[5];
+
+ void foo();
+};
+
+void S::foo() {
+#pragma acc parallel private(ThisMember, this->ThisMemberArray[1])
+ while(true);
+
+#pragma acc parallel private(ThisMemberArray[1:2])
+ while(true);
+
+#pragma acc parallel private(this)
+ while(true);
+
+#pragma acc parallel private(ThisMember, this->ThisMember)
+ while(true);
+}
+
+void Inst() {
+ static constexpr int NTTP_REFed = 1;
+ int i;
+ int Arr[5];
+ Complete C;
+ TemplUses(i, Arr, C); // #TEMPL_USES_INST
+ NTTP<5, NTTP_REFed>(); // #NTTP_INST
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
new file mode 100644
index 00000000000000..5d10106724b907
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
@@ -0,0 +1,552 @@
+// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
+
+int Global;
+short GlobalArray[5];
+
+void NormalUses(float *PointerParam) {
+ // CHECK: FunctionDecl{{.*}}NormalUses
+ // CHECK: ParmVarDecl
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel private(Global, GlobalArray[2])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'short' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'short *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 2
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(GlobalArray, PointerParam[Global])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(GlobalArray) private(PointerParam[Global])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(GlobalArray, PointerParam[Global : Global])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}} 'PointerParam' 'float *'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+
+// This example is an error typically, but we want to make sure we're properly
+// capturing NTTPs until instantiation time.
+template<unsigned I>
+void UnInstTempl() {
+ // CHECK-NEXT: FunctionTemplateDecl{{.*}} UnInstTempl
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}}referenced 'unsigned int' depth 0 index 0 I
+ // CHECK-NEXT: FunctionDecl{{.*}} UnInstTempl 'void ()'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc parallel private(I)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'unsigned int' NonTypeTemplateParm{{.*}}'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+
+template<auto &NTTP, typename T, typename U>
+void TemplUses(T t, U u, T*PointerParam) {
+ // CHECK-NEXT: FunctionTemplateDecl
+ // CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}referenced 'auto &' depth 0 index 0 NTTP
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 1 T
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 2 U
+ // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void (T, U, T *)'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced t 'T'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced u 'U'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced PointerParam 'T *'
+ // CHECK-NEXT: CompoundStmt
+
+
+#pragma acc parallel private(GlobalArray, PointerParam[Global])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'T' lvalue
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T *' lvalue ParmVar{{.*}}'PointerParam' 'T *'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(t, u)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(t) private(u)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(t) private(NTTP, u)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(u[0])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'<dependent type>' lvalue
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(u[0:t])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}}EndMarker
+ int EndMarker;
+
+ // Check the instantiated versions of the above.
+ // CHECK-NEXT: FunctionDecl{{.*}} used TemplUses 'void (int, int *, int *)' implicit_instantiation
+ // CHECK-NEXT: TemplateArgument decl
+ // CHECK-NEXT: Var{{.*}} 'CEVar' 'const unsigned int'
+ // CHECK-NEXT: TemplateArgument type 'int'
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: TemplateArgument type 'int[1]'
+ // CHECK-NEXT: ConstantArrayType{{.*}} 'int[1]' 1
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used t 'int'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used u 'int *'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used PointerParam 'int *'
+ // CHECK-NEXT: CompoundStmt
+
+// #pragma acc parallel private(GlobalArray, PointerParam[Global])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'int' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}}'PointerParam' 'int *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+// #pragma acc parallel private(t, u)
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+// #pragma acc parallel private(t) private(u)
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+// #pragma acc parallel private(t) private(NTTP, u)
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP
+ // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+// #pragma acc parallel private(u[0])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}}'int' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+// #pragma acc parallel private(u[0:t])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}}EndMarker
+}
+
+struct S {
+ // CHECK-NEXT: CXXRecordDecl{{.*}} struct S definition
+ // CHECK: CXXRecordDecl{{.*}} implicit struct S
+ int ThisMember;
+ // CHECK-NEXT: FieldDecl{{.*}} ThisMember 'int'
+ int ThisMemberArray[5];
+ // CHECK-NEXT: FieldDecl{{.*}} ThisMemberArray 'int[5]'
+
+ void foo();
+ // CHECK-NEXT: CXXMethodDecl{{.*}} foo 'void ()'
+
+ template<typename T>
+ void bar() {
+ // CHECK-NEXT: FunctionTemplateDecl{{.*}}bar
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 0 T
+ // CHECK-NEXT: CXXMethodDecl{{.*}} bar 'void ()' implicit-inline
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel private(ThisMember, this->ThisMemberArray[1])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ThisMemberArray[1:2])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(this)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+ // Check Instantiations:
+ // CHECK-NEXT: CXXMethodDecl{{.*}} used bar 'void ()' implicit_instantiation implicit-inline
+ // CHECK-NEXT: TemplateArgument type 'int'
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: CompoundStmt
+
+// #pragma acc parallel private(ThisMember, this->ThisMemberArray[1])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+// #pragma acc parallel private(ThisMemberArray[1:2])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+// #pragma acc parallel private(this)
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+};
+
+void S::foo() {
+ // CHECK: CXXMethodDecl{{.*}} foo 'void ()'
+ // CHECK-NEXT: CompoundStmt
+#pragma acc parallel private(ThisMember, this->ThisMemberArray[1])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ThisMemberArray[1:2])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(this)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+
+template<typename U>
+struct STempl {
+ // CHECK-NEXT: ClassTemplateDecl{{.*}} STempl
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}} typename depth 0 index 0 U
+ // CHECK-NEXT: CXXRecordDecl{{.*}} struct STempl definition
+ // CHECK: CXXRecordDecl{{.*}} implicit struct STempl
+ U ThisMember;
+ // CHECK-NEXT: FieldDecl{{.*}} ThisMember 'U'
+ U ThisMemberArray[5];
+ // CHECK-NEXT: FieldDecl{{.*}} ThisMemberArray 'U[5]'
+
+ void foo() {
+ // CHECK-NEXT: CXXMethodDecl {{.*}} foo 'void ()'
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel private(ThisMember, this->ThisMemberArray[1])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: MemberExpr{{.*}} 'U' lvalue ->ThisMember
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}} '<dependent type>' lvalue
+ // CHECK-NEXT: CXXDependentScopeMemberExpr{{.*}} '<dependent type>' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ThisMemberArray[1:2])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: MemberExpr{{.*}} 'U[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(this)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+
+ template<typename T>
+ void bar() {
+ // CHECK-NEXT: FunctionTemplateDecl{{.*}} bar
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}} typename depth 1 index 0 T
+ // CHECK-NEXT: CXXMethodDecl{{.*}} bar 'void ()'
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel private(ThisMember, this->ThisMemberArray[1])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: MemberExpr{{.*}} 'U' lvalue ->ThisMember
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}} '<dependent type>' lvalue
+ // CHECK-NEXT: CXXDependentScopeMemberExpr{{.*}} '<dependent type>' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ThisMemberArray[1:2])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: MemberExpr{{.*}} 'U[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(this)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+
+// Instantiation of the class template.
+
+// CHECK-NEXT: ClassTemplateSpecializationDecl{{.*}}struct STempl
+// CHECK: TemplateArgument type 'int'
+// CHECK-NEXT: BuiltinType {{.*}}'int'
+// CHECK-NEXT: CXXRecordDecl{{.*}} struct STempl
+// CHECK-NEXT: FieldDecl{{.*}}ThisMember 'int'
+// CHECK-NEXT: FieldDecl{{.*}} ThisMemberArray 'int[5]'
+
+// CHECK-NEXT: CXXMethodDecl{{.*}} foo 'void ()'
+// CHECK-NEXT: FunctionTemplateDecl{{.*}}bar
+// CHECK-NEXT: TemplateTypeParmDecl{{.*}} typename depth 0 index 0 T
+// CHECK-NEXT: CXXMethodDecl{{.*}}bar 'void ()'
+// CHECK-NEXT: CXXMethodDecl{{.*}}bar 'void ()'
+// CHECK-NEXT: TemplateArgument type 'int'
+// CHECK-NEXT: BuiltinType{{.*}} 'int'
+// CHECK-NEXT: CompoundStmt
+
+//#pragma acc parallel private(ThisMember, this->ThisMemberArray[1])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' implicit this
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(ThisMemberArray[1:2])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' implicit this
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
+ // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(this)
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' this
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+};
+
+void Inst() {
+ static constexpr unsigned CEVar = 1;
+ int i;
+ int Arr[5];
+ TemplUses<CEVar, int, int[1]>({}, {}, &i);
+
+ S s;
+ s.bar<int>();
+ STempl<int> stempl;
+ stempl.bar<int>();
+}
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index 398a11a5703558..eb0ba09c5b9116 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2782,6 +2782,11 @@ class OpenACCClauseEnqueue : public OpenACCClauseVisitor<OpenACCClauseEnqueue> {
public:
OpenACCClauseEnqueue(EnqueueVisitor &V) : Visitor(V) {}
+ void VisitVarList(const OpenACCClauseWithVarList &C) {
+ for (Expr *Var : C.getVarList())
+ Visitor.AddStmt(Var);
+ }
+
#define VISIT_CLAUSE(CLAUSE_NAME) \
void Visit##CLAUSE_NAME##Clause(const OpenACC##CLAUSE_NAME##Clause &C);
#include "clang/Basic/OpenACCClauses.def"
@@ -2807,6 +2812,10 @@ void OpenACCClauseEnqueue::VisitNumGangsClause(const OpenACCNumGangsClause &C) {
for (Expr *IE : C.getIntExprs())
Visitor.AddStmt(IE);
}
+
+void OpenACCClauseEnqueue::VisitPrivateClause(const OpenACCPrivateClause &C) {
+ VisitVarList(C);
+}
} // namespace
void EnqueueVisitor::EnqueueChildren(const OpenACCClause *C) {
>From 11a38905cc96789fd1bc2f70cb2e59526b37e68b Mon Sep 17 00:00:00 2001
From: erichkeane <ekeane at nvidia.com>
Date: Tue, 30 Apr 2024 06:16:34 -0700
Subject: [PATCH 2/2] Changes from code review
---
clang/include/clang/Sema/SemaOpenACC.h | 4 +---
clang/lib/Parse/ParseOpenACC.cpp | 6 ++----
clang/test/SemaOpenACC/compute-construct-private-clause.c | 7 +++++++
3 files changed, 10 insertions(+), 7 deletions(-)
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index 9c4ca40e34c5c4..edb0cbb7c5d552 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -116,9 +116,7 @@ class SemaOpenACC : public SemaBase {
return const_cast<OpenACCParsedClause *>(this)->getIntExprs();
}
- // Non-const version that permits modifying of the VarList for the purposes
- // of Sema enforcement.
- SmallVector<Expr *> &getVarList() {
+ ArrayRef<Expr *> getVarList() {
assert(ClauseKind == OpenACCClauseKind::Private &&
"Parsed clause kind does not have a var-list");
return std::get<VarListDetails>(Details).VarList;
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index a12ffad699755f..2d1ec6539b2fd1 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -934,11 +934,9 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
case OpenACCClauseKind::UseDevice:
ParseOpenACCVarList();
break;
- case OpenACCClauseKind::Private: {
- llvm::SmallVector<Expr *> Vars = ParseOpenACCVarList();
- ParsedClause.setVarListDetails(std::move(Vars));
+ case OpenACCClauseKind::Private:
+ ParsedClause.setVarListDetails(ParseOpenACCVarList());
break;
- }
case OpenACCClauseKind::Collapse: {
tryParseAndConsumeSpecialTokenKind(*this, OpenACCSpecialTokenKind::Force,
ClauseKind);
diff --git a/clang/test/SemaOpenACC/compute-construct-private-clause.c b/clang/test/SemaOpenACC/compute-construct-private-clause.c
index 959e1175b5e156..15775279fc8690 100644
--- a/clang/test/SemaOpenACC/compute-construct-private-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-private-clause.c
@@ -135,4 +135,11 @@ void uses(int IntParam, void *PointerParam, float ArrayParam[5], Complete Compos
while(1);
#pragma acc parallel private(ArrayParam[2])
while(1);
+
+ // expected-error at +1{{OpenACC sub-array is not allowed here}}
+#pragma acc parallel private((float*)ArrayParam[2:5])
+ while(1);
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel private((float)ArrayParam[2])
+ while(1);
}
More information about the cfe-commits
mailing list