[clang] 2c2accb - [OpenACC] Enable 'self' sema for 'update' construct
via cfe-commits
cfe-commits at lists.llvm.org
Wed Jan 8 13:19:52 PST 2025
Author: erichkeane
Date: 2025-01-08T13:19:33-08:00
New Revision: 2c2accbcc6b0f132182a35b65ac76c038912cd1e
URL: https://github.com/llvm/llvm-project/commit/2c2accbcc6b0f132182a35b65ac76c038912cd1e
DIFF: https://github.com/llvm/llvm-project/commit/2c2accbcc6b0f132182a35b65ac76c038912cd1e.diff
LOG: [OpenACC] Enable 'self' sema for 'update' construct
The 'self' clause is an unfortunately difficult one, as it has a
significantly different meaning between 'update' and the other
constructs. This patch introduces a way for the 'self' clause to work
as both. I considered making this two separate AST nodes (one for
'self' on 'update' and one for the others), however this makes the
automated macros/etc for supporting a clause break.
Instead, 'self' has the ability to act as either a condition or as a
var-list clause. As this is the only one of its kind, it is implemented
all within it. If in the future we have more that work like this, we
should consider rewriting a lot of the macros that we use to make
clauses work, and make them separate ast nodes.
Added:
Modified:
clang/include/clang/AST/OpenACCClause.h
clang/include/clang/Sema/SemaOpenACC.h
clang/lib/AST/OpenACCClause.cpp
clang/lib/AST/StmtProfile.cpp
clang/lib/Parse/ParseOpenACC.cpp
clang/lib/Sema/SemaOpenACC.cpp
clang/lib/Sema/TreeTransform.h
clang/lib/Serialization/ASTReader.cpp
clang/lib/Serialization/ASTWriter.cpp
clang/test/AST/ast-print-openacc-update-construct.cpp
clang/test/ParserOpenACC/parse-clauses.c
clang/test/SemaOpenACC/combined-construct-self-ast.cpp
clang/test/SemaOpenACC/compute-construct-clause-ast.cpp
clang/test/SemaOpenACC/update-construct-ast.cpp
clang/test/SemaOpenACC/update-construct.cpp
clang/tools/libclang/CIndex.cpp
Removed:
################################################################################
diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h
index adc5e48583d003..4e4dd3447926ee 100644
--- a/clang/include/clang/AST/OpenACCClause.h
+++ b/clang/include/clang/AST/OpenACCClause.h
@@ -327,18 +327,89 @@ class OpenACCIfClause : public OpenACCClauseWithCondition {
SourceLocation EndLoc);
};
-/// A 'self' clause, which has an optional condition expression.
-class OpenACCSelfClause : public OpenACCClauseWithCondition {
+/// A 'self' clause, which has an optional condition expression, or, in the
+/// event of an 'update' directive, contains a 'VarList'.
+class OpenACCSelfClause final
+ : public OpenACCClauseWithParams,
+ private llvm::TrailingObjects<OpenACCSelfClause, Expr *> {
+ friend TrailingObjects;
+ // Holds whether this HAS a condition expression. Lacks a value if this is NOT
+ // a condition-expr self clause.
+ std::optional<bool> HasConditionExpr;
+ // Holds the number of stored expressions. In the case of a condition-expr
+ // self clause, this is expected to be ONE (and there to be 1 trailing
+ // object), whether or not that is null.
+ unsigned NumExprs;
+
OpenACCSelfClause(SourceLocation BeginLoc, SourceLocation LParenLoc,
Expr *ConditionExpr, SourceLocation EndLoc);
+ OpenACCSelfClause(SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc);
+
+ // Intentionally internal, meant to be an implementation detail of everything
+ // else. All non-internal uses should go through getConditionExpr/getVarList.
+ llvm::ArrayRef<Expr *> getExprs() const {
+ return {getTrailingObjects<Expr *>(), NumExprs};
+ }
public:
static bool classof(const OpenACCClause *C) {
return C->getClauseKind() == OpenACCClauseKind::Self;
}
+
+ bool isConditionExprClause() const { return HasConditionExpr.has_value(); }
+
+ bool hasConditionExpr() const {
+ assert(HasConditionExpr.has_value() &&
+ "VarList Self Clause asked about condition expression");
+ return *HasConditionExpr;
+ }
+
+ const Expr *getConditionExpr() const {
+ assert(HasConditionExpr.has_value() &&
+ "VarList Self Clause asked about condition expression");
+ assert(getExprs().size() == 1 &&
+ "ConditionExpr Self Clause with too many Exprs");
+ return getExprs()[0];
+ }
+
+ Expr *getConditionExpr() {
+ assert(HasConditionExpr.has_value() &&
+ "VarList Self Clause asked about condition expression");
+ assert(getExprs().size() == 1 &&
+ "ConditionExpr Self Clause with too many Exprs");
+ return getExprs()[0];
+ }
+
+ ArrayRef<Expr *> getVarList() {
+ assert(!HasConditionExpr.has_value() &&
+ "Condition Expr self clause asked about var list");
+ return getExprs();
+ }
+ ArrayRef<Expr *> getVarList() const {
+ assert(!HasConditionExpr.has_value() &&
+ "Condition Expr self clause asked about var list");
+ return getExprs();
+ }
+
+ child_range children() {
+ return child_range(
+ reinterpret_cast<Stmt **>(getTrailingObjects<Expr *>()),
+ reinterpret_cast<Stmt **>(getTrailingObjects<Expr *>() + NumExprs));
+ }
+
+ const_child_range children() const {
+ child_range Children = const_cast<OpenACCSelfClause *>(this)->children();
+ return const_child_range(Children.begin(), Children.end());
+ }
+
static OpenACCSelfClause *Create(const ASTContext &C, SourceLocation BeginLoc,
SourceLocation LParenLoc,
Expr *ConditionExpr, SourceLocation EndLoc);
+ static OpenACCSelfClause *Create(const ASTContext &C, SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ ArrayRef<Expr *> ConditionExpr,
+ SourceLocation EndLoc);
};
/// Represents a clause that has one or more expressions associated with it.
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index 03abf4ab2cec87..0f86d46bc98025 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -409,6 +409,8 @@ class SemaOpenACC : public SemaBase {
ClauseKind == OpenACCClauseKind::Detach ||
ClauseKind == OpenACCClauseKind::DevicePtr ||
ClauseKind == OpenACCClauseKind::Reduction ||
+ (ClauseKind == OpenACCClauseKind::Self &&
+ DirKind == OpenACCDirectiveKind::Update) ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
"Parsed clause kind does not have a var-list");
@@ -551,6 +553,8 @@ class SemaOpenACC : public SemaBase {
ClauseKind == OpenACCClauseKind::UseDevice ||
ClauseKind == OpenACCClauseKind::Detach ||
ClauseKind == OpenACCClauseKind::DevicePtr ||
+ (ClauseKind == OpenACCClauseKind::Self &&
+ DirKind == OpenACCDirectiveKind::Update) ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
"Parsed clause kind does not have a var-list");
assert((!IsReadOnly || ClauseKind == OpenACCClauseKind::CopyIn ||
@@ -590,6 +594,8 @@ class SemaOpenACC : public SemaBase {
ClauseKind == OpenACCClauseKind::UseDevice ||
ClauseKind == OpenACCClauseKind::Detach ||
ClauseKind == OpenACCClauseKind::DevicePtr ||
+ (ClauseKind == OpenACCClauseKind::Self &&
+ DirKind == OpenACCDirectiveKind::Update) ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
"Parsed clause kind does not have a var-list");
assert((!IsReadOnly || ClauseKind == OpenACCClauseKind::CopyIn ||
diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp
index 76fea1fd47d217..da63b471d98565 100644
--- a/clang/lib/AST/OpenACCClause.cpp
+++ b/clang/lib/AST/OpenACCClause.cpp
@@ -20,7 +20,7 @@ using namespace clang;
bool OpenACCClauseWithParams::classof(const OpenACCClause *C) {
return OpenACCDeviceTypeClause::classof(C) ||
OpenACCClauseWithCondition::classof(C) ||
- OpenACCClauseWithExprs::classof(C);
+ OpenACCClauseWithExprs::classof(C) || OpenACCSelfClause::classof(C);
}
bool OpenACCClauseWithExprs::classof(const OpenACCClause *C) {
return OpenACCWaitClause::classof(C) || OpenACCNumGangsClause::classof(C) ||
@@ -41,7 +41,7 @@ bool OpenACCClauseWithVarList::classof(const OpenACCClause *C) {
OpenACCReductionClause::classof(C) || OpenACCCreateClause::classof(C);
}
bool OpenACCClauseWithCondition::classof(const OpenACCClause *C) {
- return OpenACCIfClause::classof(C) || OpenACCSelfClause::classof(C);
+ return OpenACCIfClause::classof(C);
}
bool OpenACCClauseWithSingleIntExpr::classof(const OpenACCClause *C) {
return OpenACCNumWorkersClause::classof(C) ||
@@ -87,19 +87,43 @@ OpenACCSelfClause *OpenACCSelfClause::Create(const ASTContext &C,
SourceLocation LParenLoc,
Expr *ConditionExpr,
SourceLocation EndLoc) {
- void *Mem = C.Allocate(sizeof(OpenACCIfClause), alignof(OpenACCIfClause));
+ void *Mem = C.Allocate(OpenACCSelfClause::totalSizeToAlloc<Expr *>(1));
return new (Mem)
OpenACCSelfClause(BeginLoc, LParenLoc, ConditionExpr, EndLoc);
}
+OpenACCSelfClause *OpenACCSelfClause::Create(const ASTContext &C,
+ SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList,
+ SourceLocation EndLoc) {
+ void *Mem =
+ C.Allocate(OpenACCSelfClause::totalSizeToAlloc<Expr *>(VarList.size()));
+ return new (Mem) OpenACCSelfClause(BeginLoc, LParenLoc, VarList, EndLoc);
+}
+
+OpenACCSelfClause::OpenACCSelfClause(SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ llvm::ArrayRef<Expr *> VarList,
+ SourceLocation EndLoc)
+ : OpenACCClauseWithParams(OpenACCClauseKind::Self, BeginLoc, LParenLoc,
+ EndLoc),
+ HasConditionExpr(std::nullopt), NumExprs(VarList.size()) {
+ std::uninitialized_copy(VarList.begin(), VarList.end(),
+ getTrailingObjects<Expr *>());
+}
+
OpenACCSelfClause::OpenACCSelfClause(SourceLocation BeginLoc,
SourceLocation LParenLoc,
Expr *ConditionExpr, SourceLocation EndLoc)
- : OpenACCClauseWithCondition(OpenACCClauseKind::Self, BeginLoc, LParenLoc,
- ConditionExpr, EndLoc) {
+ : OpenACCClauseWithParams(OpenACCClauseKind::Self, BeginLoc, LParenLoc,
+ EndLoc),
+ HasConditionExpr(ConditionExpr != nullptr), NumExprs(1) {
assert((!ConditionExpr || ConditionExpr->isInstantiationDependent() ||
ConditionExpr->getType()->isScalarType()) &&
"Condition expression type not scalar/dependent");
+ std::uninitialized_copy(&ConditionExpr, &ConditionExpr + 1,
+ getTrailingObjects<Expr *>());
}
OpenACCClause::child_range OpenACCClause::children() {
@@ -555,9 +579,17 @@ void OpenACCClausePrinter::VisitIfClause(const OpenACCIfClause &C) {
void OpenACCClausePrinter::VisitSelfClause(const OpenACCSelfClause &C) {
OS << "self";
- if (const Expr *CondExpr = C.getConditionExpr()) {
+
+ if (C.isConditionExprClause()) {
+ if (const Expr *CondExpr = C.getConditionExpr()) {
+ OS << "(";
+ printExpr(CondExpr);
+ OS << ")";
+ }
+ } else {
OS << "(";
- printExpr(CondExpr);
+ llvm::interleaveComma(C.getVarList(), OS,
+ [&](const Expr *E) { printExpr(E); });
OS << ")";
}
}
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index b68c83f99550b3..cd91a7900538ba 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -2555,8 +2555,13 @@ void OpenACCClauseProfiler::VisitCreateClause(
}
void OpenACCClauseProfiler::VisitSelfClause(const OpenACCSelfClause &Clause) {
- if (Clause.hasConditionExpr())
- Profiler.VisitStmt(Clause.getConditionExpr());
+ if (Clause.isConditionExprClause()) {
+ if (Clause.hasConditionExpr())
+ Profiler.VisitStmt(Clause.getConditionExpr());
+ } else {
+ for (auto *E : Clause.getVarList())
+ Profiler.VisitStmt(E);
+ }
}
void OpenACCClauseProfiler::VisitFinalizeClause(
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index a9deae74cf27c6..c79ba97a200778 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -1003,7 +1003,9 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
// the 'update' clause, so we have to handle it here. U se an assert to
// make sure we get the right
diff erentiator.
assert(DirKind == OpenACCDirectiveKind::Update);
- [[fallthrough]];
+ ParsedClause.setVarListDetails(ParseOpenACCVarList(ClauseKind),
+ /*IsReadOnly=*/false, /*IsZero=*/false);
+ break;
case OpenACCClauseKind::Device:
case OpenACCClauseKind::DeviceResident:
case OpenACCClauseKind::Host:
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 1edff48331cd6d..51a95f99f0624a 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -736,14 +736,14 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitIfClause(
// isn't really much to do here.
// If the 'if' clause is true, it makes the 'self' clause have no effect,
- // diagnose that here.
- // TODO OpenACC: When we add these two to other constructs, we might not
- // want to warn on this (for example, 'update').
- const auto *Itr =
- llvm::find_if(ExistingClauses, llvm::IsaPred<OpenACCSelfClause>);
- if (Itr != ExistingClauses.end()) {
- SemaRef.Diag(Clause.getBeginLoc(), diag::warn_acc_if_self_conflict);
- SemaRef.Diag((*Itr)->getBeginLoc(), diag::note_acc_previous_clause_here);
+ // diagnose that here. This only applies on compute/combined constructs.
+ if (Clause.getDirectiveKind() != OpenACCDirectiveKind::Update) {
+ const auto *Itr =
+ llvm::find_if(ExistingClauses, llvm::IsaPred<OpenACCSelfClause>);
+ if (Itr != ExistingClauses.end()) {
+ SemaRef.Diag(Clause.getBeginLoc(), diag::warn_acc_if_self_conflict);
+ SemaRef.Diag((*Itr)->getBeginLoc(), diag::note_acc_previous_clause_here);
+ }
}
return OpenACCIfClause::Create(Ctx, Clause.getBeginLoc(),
@@ -753,16 +753,6 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitIfClause(
OpenACCClause *SemaOpenACCClauseVisitor::VisitSelfClause(
SemaOpenACC::OpenACCParsedClause &Clause) {
- // 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 (!isDirectiveKindImplemented(Clause.getDirectiveKind()))
- return isNotImplemented();
-
- // TODO OpenACC: When we implement this for 'update', this takes a
- // 'var-list' instead of a condition expression, so semantics/handling has
- // to happen
diff erently here.
-
// There is no prose in the standard that says duplicates aren't allowed,
// but this diagnostic is present in other compilers, as well as makes
// sense.
@@ -770,9 +760,12 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitSelfClause(
return nullptr;
// If the 'if' clause is true, it makes the 'self' clause have no effect,
- // diagnose that here.
- // TODO OpenACC: When we add these two to other constructs, we might not
- // want to warn on this (for example, 'update').
+ // diagnose that here. This only applies on compute/combined constructs.
+ if (Clause.getDirectiveKind() == OpenACCDirectiveKind::Update)
+ return OpenACCSelfClause::Create(Ctx, Clause.getBeginLoc(),
+ Clause.getLParenLoc(), Clause.getVarList(),
+ Clause.getEndLoc());
+
const auto *Itr =
llvm::find_if(ExistingClauses, llvm::IsaPred<OpenACCIfClause>);
if (Itr != ExistingClauses.end()) {
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index 15ea8c7128f38d..d00ad5a35e8235 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -11647,22 +11647,48 @@ template <typename Derived>
void OpenACCClauseTransform<Derived>::VisitSelfClause(
const OpenACCSelfClause &C) {
- if (C.hasConditionExpr()) {
- Expr *Cond = const_cast<Expr *>(C.getConditionExpr());
- Sema::ConditionResult Res =
- Self.TransformCondition(Cond->getExprLoc(), /*Var=*/nullptr, Cond,
- Sema::ConditionKind::Boolean);
+ // If this is an 'update' 'self' clause, this is actually a var list instead.
+ if (ParsedClause.getDirectiveKind() == OpenACCDirectiveKind::Update) {
+ llvm::SmallVector<Expr *> InstantiatedVarList;
+ for (Expr *CurVar : C.getVarList()) {
+ ExprResult Res = Self.TransformExpr(CurVar);
- if (Res.isInvalid() || !Res.get().second)
- return;
+ if (!Res.isUsable())
+ continue;
- ParsedClause.setConditionDetails(Res.get().second);
- }
+ Res = Self.getSema().OpenACC().ActOnVar(ParsedClause.getClauseKind(),
+ Res.get());
- NewClause = OpenACCSelfClause::Create(
- Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
- ParsedClause.getLParenLoc(), ParsedClause.getConditionExpr(),
- ParsedClause.getEndLoc());
+ if (Res.isUsable())
+ InstantiatedVarList.push_back(Res.get());
+ }
+
+ ParsedClause.setVarListDetails(InstantiatedVarList,
+ /*IsReadOnly=*/false, /*IsZero=*/false);
+
+ NewClause = OpenACCSelfClause::Create(
+ Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
+ ParsedClause.getLParenLoc(), ParsedClause.getVarList(),
+ ParsedClause.getEndLoc());
+ } else {
+
+ if (C.hasConditionExpr()) {
+ Expr *Cond = const_cast<Expr *>(C.getConditionExpr());
+ Sema::ConditionResult Res =
+ Self.TransformCondition(Cond->getExprLoc(), /*Var=*/nullptr, Cond,
+ Sema::ConditionKind::Boolean);
+
+ if (Res.isInvalid() || !Res.get().second)
+ return;
+
+ ParsedClause.setConditionDetails(Res.get().second);
+ }
+
+ NewClause = OpenACCSelfClause::Create(
+ Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
+ ParsedClause.getLParenLoc(), ParsedClause.getConditionExpr(),
+ ParsedClause.getEndLoc());
+ }
}
template <typename Derived>
diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp
index 0c82e540047f8d..0368990ca150df 100644
--- a/clang/lib/Serialization/ASTReader.cpp
+++ b/clang/lib/Serialization/ASTReader.cpp
@@ -12387,9 +12387,18 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
}
case OpenACCClauseKind::Self: {
SourceLocation LParenLoc = readSourceLocation();
- Expr *CondExpr = readBool() ? readSubExpr() : nullptr;
- return OpenACCSelfClause::Create(getContext(), BeginLoc, LParenLoc,
- CondExpr, EndLoc);
+ bool isConditionExprClause = readBool();
+ if (isConditionExprClause) {
+ Expr *CondExpr = readBool() ? readSubExpr() : nullptr;
+ return OpenACCSelfClause::Create(getContext(), BeginLoc, LParenLoc,
+ CondExpr, EndLoc);
+ }
+ unsigned NumVars = readInt();
+ llvm::SmallVector<Expr *> VarList;
+ for (unsigned I = 0; I < NumVars; ++I)
+ VarList.push_back(readSubExpr());
+ return OpenACCSelfClause::Create(getContext(), BeginLoc, LParenLoc, VarList,
+ EndLoc);
}
case OpenACCClauseKind::NumGangs: {
SourceLocation LParenLoc = readSourceLocation();
diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp
index 39f8b0fd5ba0f9..8d9396e28ed50a 100644
--- a/clang/lib/Serialization/ASTWriter.cpp
+++ b/clang/lib/Serialization/ASTWriter.cpp
@@ -8321,9 +8321,16 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
case OpenACCClauseKind::Self: {
const auto *SC = cast<OpenACCSelfClause>(C);
writeSourceLocation(SC->getLParenLoc());
- writeBool(SC->hasConditionExpr());
- if (SC->hasConditionExpr())
- AddStmt(const_cast<Expr*>(SC->getConditionExpr()));
+ writeBool(SC->isConditionExprClause());
+ if (SC->isConditionExprClause()) {
+ writeBool(SC->hasConditionExpr());
+ if (SC->hasConditionExpr())
+ AddStmt(const_cast<Expr *>(SC->getConditionExpr()));
+ } else {
+ writeUInt32(SC->getVarList().size());
+ for (Expr *E : SC->getVarList())
+ AddStmt(E);
+ }
return;
}
case OpenACCClauseKind::NumGangs: {
diff --git a/clang/test/AST/ast-print-openacc-update-construct.cpp b/clang/test/AST/ast-print-openacc-update-construct.cpp
index ce83bcad003a2a..a7f5b2a42285be 100644
--- a/clang/test/AST/ast-print-openacc-update-construct.cpp
+++ b/clang/test/AST/ast-print-openacc-update-construct.cpp
@@ -35,4 +35,7 @@ void uses(bool cond) {
// CHECK: #pragma acc update device_type(J) dtype(K)
#pragma acc update device_type(J) dtype(K)
+
+// CHECK: #pragma acc update self(I, iPtr, array, array[1], array[1:2])
+#pragma acc update self(I, iPtr, array, array[1], array[1:2])
}
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index 9b88c147d0faa2..73a09697710f98 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -347,14 +347,12 @@ void SelfUpdate() {
#pragma acc update self
for(int i = 0; i < 5;++i) {}
- // expected-error at +4{{use of undeclared identifier 'zero'}}
- // expected-error at +3{{expected ','}}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented, clause ignored}}
+ // expected-error at +3{{use of undeclared identifier 'zero'}}
+ // expected-error at +2{{expected ','}}
+ // expected-error at +1{{expected expression}}
#pragma acc update self(zero : s.array[s.value : 5], s.value), if_present
for(int i = 0; i < 5;++i) {}
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented, clause ignored}}
#pragma acc update self(s.array[s.value : 5], s.value), if_present
for(int i = 0; i < 5;++i) {}
}
diff --git a/clang/test/SemaOpenACC/combined-construct-self-ast.cpp b/clang/test/SemaOpenACC/combined-construct-self-ast.cpp
index 3a6ba3ca6aea23..e504ea7f5a0756 100644
--- a/clang/test/SemaOpenACC/combined-construct-self-ast.cpp
+++ b/clang/test/SemaOpenACC/combined-construct-self-ast.cpp
@@ -20,6 +20,7 @@ void TemplFunc() {
for (unsigned i = 0; i < 5; ++i);
// CHECK-NEXT: OpenACCCombinedConstruct{{.*}}serial loop
// CHECK-NEXT: self clause
+ // CHECK-NEXT: <<<NULL>>
// CHECK-NEXT: ForStmt
// CHECK: NullStmt
@@ -65,6 +66,7 @@ void TemplFunc() {
//
// CHECK-NEXT: OpenACCCombinedConstruct{{.*}}serial loop
// CHECK-NEXT: self clause
+ // CHECK-NEXT: <<<NULL>>
// CHECK-NEXT: ForStmt
// CHECK: NullStmt
diff --git a/clang/test/SemaOpenACC/compute-construct-clause-ast.cpp b/clang/test/SemaOpenACC/compute-construct-clause-ast.cpp
index 69f65f4083ae7b..58c12b828439df 100644
--- a/clang/test/SemaOpenACC/compute-construct-clause-ast.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-clause-ast.cpp
@@ -197,6 +197,7 @@ void TemplFunc() {
while(true);
// CHECK-NEXT: OpenACCComputeConstruct{{.*}}serial
// CHECK-NEXT: self clause
+ // CHECK-NEXT: <<<NULL>>
// CHECK-NEXT: WhileStmt
// CHECK-NEXT: CXXBoolLiteralExpr
// CHECK-NEXT: NullStmt
@@ -393,6 +394,7 @@ void TemplFunc() {
// CHECK-NEXT: OpenACCComputeConstruct{{.*}}serial
// CHECK-NEXT: self clause
+ // CHECK-NEXT: <<<NULL>>
// CHECK-NEXT: WhileStmt
// CHECK-NEXT: CXXBoolLiteralExpr
// CHECK-NEXT: NullStmt
diff --git a/clang/test/SemaOpenACC/update-construct-ast.cpp b/clang/test/SemaOpenACC/update-construct-ast.cpp
index 114de654670d30..3638d7edafed10 100644
--- a/clang/test/SemaOpenACC/update-construct-ast.cpp
+++ b/clang/test/SemaOpenACC/update-construct-ast.cpp
@@ -10,6 +10,10 @@
int some_int();
long some_long();
+int Global;
+short GlobalArray[5];
+
+
void NormalFunc() {
// CHECK-LABEL: NormalFunc
// CHECK-NEXT: CompoundStmt
@@ -70,6 +74,21 @@ void NormalFunc() {
// CHECK-NEXT: CallExpr{{.*}}'long'
// CHECK-NEXT: ImplicitCastExpr
// CHECK-NEXT: DeclRefExpr{{.*}}'some_long' 'long ()'
+
+#pragma acc update self(Global, GlobalArray, GlobalArray[0], GlobalArray[0:1])
+ // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
+ // CHECK-NEXT: self clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'Global' 'int'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'short' lvalue
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 0
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 0
+ // CHECK-NEXT: IntegerLiteral{{.*}} 1
}
template<typename T>
@@ -124,6 +143,26 @@ void TemplFunc(T t) {
// CHECK-NEXT: DependentScopeDeclRefExpr{{.*}}'<dependent type>'
// CHECK-NEXT: NestedNameSpecifier TypeSpec 'T'
+ decltype(T::value) Local = 0, LocalArray[5] = {};
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl
+ // CHECK-NEXT: IntegerLiteral
+ // CHECK-NEXT: VarDecl
+ // CHECK-NEXT: InitListExpr
+
+#pragma acc update self(Local, LocalArray, LocalArray[0], LocalArray[0:1])
+ // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
+ // CHECK-NEXT: self clause
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'Local' 'decltype(T::value)'
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'LocalArray' 'decltype(T::value)[5]'
+ // CHECK-NEXT: ArraySubscriptExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'LocalArray' 'decltype(T::value)[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}}0
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'LocalArray' 'decltype(T::value)[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}}0
+ // CHECK-NEXT: IntegerLiteral{{.*}}1
+
// Instantiation:
// CHECK-NEXT: FunctionDecl{{.*}} TemplFunc 'void (SomeStruct)' implicit_instantiation
// CHECK-NEXT: TemplateArgument type 'SomeStruct'
@@ -194,6 +233,27 @@ void TemplFunc(T t) {
// CHECK-NEXT: ImplicitCastExpr{{.*}}'unsigned int'
// CHECK-NEXT: DeclRefExpr{{.*}}'value' 'const unsigned int'
// CHECK-NEXT: NestedNameSpecifier TypeSpec 'SomeStruct'
+
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: IntegerLiteral
+ // CHECK-NEXT: VarDecl
+ // CHECK-NEXT: InitListExpr
+ // CHECK-NEXT: array_filler
+
+ // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
+ // CHECK-NEXT: self clause
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'Local' 'decltype(SomeStruct::value)':'const unsigned int'
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'LocalArray' 'decltype(SomeStruct::value)[5]'
+ // CHECK-NEXT: ArraySubscriptExpr
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'LocalArray' 'decltype(SomeStruct::value)[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}}0
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'LocalArray' 'decltype(SomeStruct::value)[5]'
+ // CHECK-NEXT: IntegerLiteral{{.*}}1
}
struct SomeStruct{
diff --git a/clang/test/SemaOpenACC/update-construct.cpp b/clang/test/SemaOpenACC/update-construct.cpp
index 04c0aaaab99ae1..2abd7a30eda888 100644
--- a/clang/test/SemaOpenACC/update-construct.cpp
+++ b/clang/test/SemaOpenACC/update-construct.cpp
@@ -4,28 +4,20 @@ struct NotConvertible{} NC;
int getI();
void uses() {
int Var;
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update async self(Var)
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update wait self(Var)
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update self(Var) device_type(I)
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update if(true) self(Var)
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update if_present self(Var)
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update self(Var)
// expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
#pragma acc update host(Var)
// expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
#pragma acc update device(Var)
- // expected-warning at +3{{OpenACC clause 'self' not yet implemented}}
// expected-error at +2{{OpenACC clause 'if' may not follow a 'device_type' clause in a 'update' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc update self(Var) device_type(I) if(true)
- // expected-warning at +3{{OpenACC clause 'self' not yet implemented}}
// expected-error at +2{{OpenACC clause 'if_present' may not follow a 'device_type' clause in a 'update' construct}}
// expected-note at +1{{previous clause is here}}
#pragma acc update self(Var) device_type(I) if_present
@@ -39,12 +31,9 @@ void uses() {
// expected-note at +1{{previous clause is here}}
#pragma acc update device_type(I) device(Var)
// These 2 are OK.
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update self(Var) device_type(I) async
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update self(Var) device_type(I) wait
// Unless otherwise specified, we assume 'device_type' can happen after itself.
- // expected-warning at +1{{OpenACC clause 'self' not yet implemented}}
#pragma acc update self(Var) device_type(I) device_type(I)
// TODO: OpenACC: These should diagnose because there isn't at least 1 of
@@ -128,3 +117,51 @@ void uses() {
// expected-error at +1{{OpenACC clause 'wait' requires expression of integer type ('struct NotConvertible' invalid)}}
#pragma acc update wait(devnum:arr : queues: arr, NC, 5)
}
+
+struct SomeS {
+ int Array[5];
+ int MemberOfComp;
+};
+
+template<typename I, typename T>
+void varlist_restrictions_templ() {
+ I iArray[5];
+ T Single;
+ T Array[5];
+
+ // Members of a subarray of struct or class type may not appear, but others
+ // are permitted to.
+#pragma acc update self(iArray[0:1])
+
+#pragma acc update self(Array[0:1])
+
+ // expected-error at +1{{OpenACC sub-array is not allowed here}}
+#pragma acc update self(Array[0:1].MemberOfComp)
+}
+
+void varlist_restrictions() {
+ varlist_restrictions_templ<int, SomeS>();// expected-note{{in instantiation of}}
+ int iArray[5];
+ SomeS Single;
+ SomeS Array[5];
+
+ int LocalInt;
+ int *LocalPtr;
+
+#pragma acc update self(LocalInt, LocalPtr, Single)
+
+#pragma acc update self(Single.MemberOfComp)
+
+#pragma acc update self(Single.Array[0:1])
+
+
+ // Members of a subarray of struct or class type may not appear, but others
+ // are permitted to.
+#pragma acc update self(iArray[0:1])
+
+#pragma acc update self(Array[0:1])
+
+ // expected-error at +1{{OpenACC sub-array is not allowed here}}
+#pragma acc update self(Array[0:1].MemberOfComp)
+}
+
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index 4114d9a37f1ecd..5e51fc4e2f66c2 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2839,8 +2839,13 @@ void OpenACCClauseEnqueue::VisitIfClause(const OpenACCIfClause &C) {
Visitor.AddStmt(C.getConditionExpr());
}
void OpenACCClauseEnqueue::VisitSelfClause(const OpenACCSelfClause &C) {
- if (C.hasConditionExpr())
- Visitor.AddStmt(C.getConditionExpr());
+ if (C.isConditionExprClause()) {
+ if (C.hasConditionExpr())
+ Visitor.AddStmt(C.getConditionExpr());
+ } else {
+ for (Expr *Var : C.getVarList())
+ Visitor.AddStmt(Var);
+ }
}
void OpenACCClauseEnqueue::VisitNumWorkersClause(
const OpenACCNumWorkersClause &C) {
More information about the cfe-commits
mailing list