[clang] be32621 - [OpenACC] Implement 'device' and 'host' clauses for 'update'
via cfe-commits
cfe-commits at lists.llvm.org
Thu Jan 9 09:29:04 PST 2025
Author: erichkeane
Date: 2025-01-09T09:28:58-08:00
New Revision: be32621ce8cbffe674c62e87c0c51c9fc4a21e5f
URL: https://github.com/llvm/llvm-project/commit/be32621ce8cbffe674c62e87c0c51c9fc4a21e5f
DIFF: https://github.com/llvm/llvm-project/commit/be32621ce8cbffe674c62e87c0c51c9fc4a21e5f.diff
LOG: [OpenACC] Implement 'device' and 'host' clauses for 'update'
These two clauses just take a 'var-list' and specify where the variables
should be copied from/to. This patch implements the AST nodes for them
and ensures they properly take a var-list.
Added:
Modified:
clang/include/clang/AST/OpenACCClause.h
clang/include/clang/Basic/OpenACCClauses.def
clang/include/clang/Sema/SemaOpenACC.h
clang/lib/AST/OpenACCClause.cpp
clang/lib/AST/StmtProfile.cpp
clang/lib/AST/TextNodeDumper.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-auto_seq_independent-clauses.c
clang/test/SemaOpenACC/combined-construct-device_type-clause.c
clang/test/SemaOpenACC/compute-construct-device_type-clause.c
clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c
clang/test/SemaOpenACC/loop-construct-device_type-clause.c
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 4e4dd3447926ee..f5be54bdada8b5 100644
--- a/clang/include/clang/AST/OpenACCClause.h
+++ b/clang/include/clang/AST/OpenACCClause.h
@@ -965,6 +965,52 @@ class OpenACCPresentClause final
Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc,
ArrayRef<Expr *> VarList, SourceLocation EndLoc);
};
+class OpenACCHostClause final
+ : public OpenACCClauseWithVarList,
+ private llvm::TrailingObjects<OpenACCHostClause, Expr *> {
+ friend TrailingObjects;
+
+ OpenACCHostClause(SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc)
+ : OpenACCClauseWithVarList(OpenACCClauseKind::Host, BeginLoc, LParenLoc,
+ EndLoc) {
+ std::uninitialized_copy(VarList.begin(), VarList.end(),
+ getTrailingObjects<Expr *>());
+ setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size()));
+ }
+
+public:
+ static bool classof(const OpenACCClause *C) {
+ return C->getClauseKind() == OpenACCClauseKind::Host;
+ }
+ static OpenACCHostClause *Create(const ASTContext &C, SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList,
+ SourceLocation EndLoc);
+};
+
+class OpenACCDeviceClause final
+ : public OpenACCClauseWithVarList,
+ private llvm::TrailingObjects<OpenACCDeviceClause, Expr *> {
+ friend TrailingObjects;
+
+ OpenACCDeviceClause(SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc)
+ : OpenACCClauseWithVarList(OpenACCClauseKind::Device, BeginLoc, LParenLoc,
+ EndLoc) {
+ std::uninitialized_copy(VarList.begin(), VarList.end(),
+ getTrailingObjects<Expr *>());
+ setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size()));
+ }
+
+public:
+ static bool classof(const OpenACCClause *C) {
+ return C->getClauseKind() == OpenACCClauseKind::Device;
+ }
+ static OpenACCDeviceClause *
+ Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc);
+};
class OpenACCCopyClause final
: public OpenACCClauseWithVarList,
diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def
index d6fb015c64913a..8b15007c85557e 100644
--- a/clang/include/clang/Basic/OpenACCClauses.def
+++ b/clang/include/clang/Basic/OpenACCClauses.def
@@ -41,6 +41,7 @@ VISIT_CLAUSE(Default)
VISIT_CLAUSE(DefaultAsync)
VISIT_CLAUSE(Delete)
VISIT_CLAUSE(Detach)
+VISIT_CLAUSE(Device)
VISIT_CLAUSE(DeviceNum)
VISIT_CLAUSE(DevicePtr)
VISIT_CLAUSE(DeviceType)
@@ -48,6 +49,7 @@ CLAUSE_ALIAS(DType, DeviceType, false)
VISIT_CLAUSE(Finalize)
VISIT_CLAUSE(FirstPrivate)
VISIT_CLAUSE(Gang)
+VISIT_CLAUSE(Host)
VISIT_CLAUSE(If)
VISIT_CLAUSE(IfPresent)
VISIT_CLAUSE(Independent)
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index 0f86d46bc98025..2e5a0ea0aaac64 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::Host ||
+ ClauseKind == OpenACCClauseKind::Device ||
(ClauseKind == OpenACCClauseKind::Self &&
DirKind == OpenACCDirectiveKind::Update) ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
@@ -553,6 +555,8 @@ class SemaOpenACC : public SemaBase {
ClauseKind == OpenACCClauseKind::UseDevice ||
ClauseKind == OpenACCClauseKind::Detach ||
ClauseKind == OpenACCClauseKind::DevicePtr ||
+ ClauseKind == OpenACCClauseKind::Host ||
+ ClauseKind == OpenACCClauseKind::Device ||
(ClauseKind == OpenACCClauseKind::Self &&
DirKind == OpenACCDirectiveKind::Update) ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
@@ -594,6 +598,8 @@ class SemaOpenACC : public SemaBase {
ClauseKind == OpenACCClauseKind::UseDevice ||
ClauseKind == OpenACCClauseKind::Detach ||
ClauseKind == OpenACCClauseKind::DevicePtr ||
+ ClauseKind == OpenACCClauseKind::Host ||
+ ClauseKind == OpenACCClauseKind::Device ||
(ClauseKind == OpenACCClauseKind::Self &&
DirKind == OpenACCDirectiveKind::Update) ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp
index da63b471d98565..aa14ab902ba661 100644
--- a/clang/lib/AST/OpenACCClause.cpp
+++ b/clang/lib/AST/OpenACCClause.cpp
@@ -38,7 +38,9 @@ bool OpenACCClauseWithVarList::classof(const OpenACCClause *C) {
OpenACCNoCreateClause::classof(C) ||
OpenACCPresentClause::classof(C) || OpenACCCopyClause::classof(C) ||
OpenACCCopyInClause::classof(C) || OpenACCCopyOutClause::classof(C) ||
- OpenACCReductionClause::classof(C) || OpenACCCreateClause::classof(C);
+ OpenACCReductionClause::classof(C) ||
+ OpenACCCreateClause::classof(C) || OpenACCDeviceClause::classof(C) ||
+ OpenACCHostClause::classof(C);
}
bool OpenACCClauseWithCondition::classof(const OpenACCClause *C) {
return OpenACCIfClause::classof(C);
@@ -406,6 +408,26 @@ OpenACCPresentClause *OpenACCPresentClause::Create(const ASTContext &C,
return new (Mem) OpenACCPresentClause(BeginLoc, LParenLoc, VarList, EndLoc);
}
+OpenACCHostClause *OpenACCHostClause::Create(const ASTContext &C,
+ SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList,
+ SourceLocation EndLoc) {
+ void *Mem =
+ C.Allocate(OpenACCHostClause::totalSizeToAlloc<Expr *>(VarList.size()));
+ return new (Mem) OpenACCHostClause(BeginLoc, LParenLoc, VarList, EndLoc);
+}
+
+OpenACCDeviceClause *OpenACCDeviceClause::Create(const ASTContext &C,
+ SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList,
+ SourceLocation EndLoc) {
+ void *Mem =
+ C.Allocate(OpenACCDeviceClause::totalSizeToAlloc<Expr *>(VarList.size()));
+ return new (Mem) OpenACCDeviceClause(BeginLoc, LParenLoc, VarList, EndLoc);
+}
+
OpenACCCopyClause *
OpenACCCopyClause::Create(const ASTContext &C, OpenACCClauseKind Spelling,
SourceLocation BeginLoc, SourceLocation LParenLoc,
@@ -711,6 +733,20 @@ void OpenACCClausePrinter::VisitPresentClause(const OpenACCPresentClause &C) {
OS << ")";
}
+void OpenACCClausePrinter::VisitHostClause(const OpenACCHostClause &C) {
+ OS << "host(";
+ llvm::interleaveComma(C.getVarList(), OS,
+ [&](const Expr *E) { printExpr(E); });
+ OS << ")";
+}
+
+void OpenACCClausePrinter::VisitDeviceClause(const OpenACCDeviceClause &C) {
+ OS << "device(";
+ llvm::interleaveComma(C.getVarList(), OS,
+ [&](const Expr *E) { printExpr(E); });
+ OS << ")";
+}
+
void OpenACCClausePrinter::VisitCopyClause(const OpenACCCopyClause &C) {
OS << C.getClauseKind() << '(';
llvm::interleaveComma(C.getVarList(), OS,
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index cd91a7900538ba..0f1ebc68a4f762 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -2554,6 +2554,15 @@ void OpenACCClauseProfiler::VisitCreateClause(
VisitClauseWithVarList(Clause);
}
+void OpenACCClauseProfiler::VisitHostClause(const OpenACCHostClause &Clause) {
+ VisitClauseWithVarList(Clause);
+}
+
+void OpenACCClauseProfiler::VisitDeviceClause(
+ const OpenACCDeviceClause &Clause) {
+ VisitClauseWithVarList(Clause);
+}
+
void OpenACCClauseProfiler::VisitSelfClause(const OpenACCSelfClause &Clause) {
if (Clause.isConditionExprClause()) {
if (Clause.hasConditionExpr())
diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp
index eedd8faad9e85f..670641242cae2f 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -408,11 +408,13 @@ void TextNodeDumper::Visit(const OpenACCClause *C) {
case OpenACCClauseKind::Copy:
case OpenACCClauseKind::PCopy:
case OpenACCClauseKind::PresentOrCopy:
+ case OpenACCClauseKind::Host:
case OpenACCClauseKind::If:
case OpenACCClauseKind::IfPresent:
case OpenACCClauseKind::Independent:
case OpenACCClauseKind::Detach:
case OpenACCClauseKind::Delete:
+ case OpenACCClauseKind::Device:
case OpenACCClauseKind::DeviceNum:
case OpenACCClauseKind::DefaultAsync:
case OpenACCClauseKind::DevicePtr:
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index c79ba97a200778..98fd61913e5a46 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -1000,15 +1000,16 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
}
case OpenACCClauseKind::Self:
// The 'self' clause is a var-list instead of a 'condition' in the case of
- // the 'update' clause, so we have to handle it here. U se an assert to
+ // the 'update' clause, so we have to handle it here. Use an assert to
// make sure we get the right
diff erentiator.
assert(DirKind == OpenACCDirectiveKind::Update);
+ [[fallthrough]];
+ case OpenACCClauseKind::Device:
+ case OpenACCClauseKind::Host:
ParsedClause.setVarListDetails(ParseOpenACCVarList(ClauseKind),
/*IsReadOnly=*/false, /*IsZero=*/false);
break;
- case OpenACCClauseKind::Device:
case OpenACCClauseKind::DeviceResident:
- case OpenACCClauseKind::Host:
case OpenACCClauseKind::Link:
ParseOpenACCVarList(ClauseKind);
break;
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 51a95f99f0624a..a2973f1f99b2cf 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -471,6 +471,22 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind,
return false;
}
}
+ case OpenACCClauseKind::Device: {
+ switch (DirectiveKind) {
+ case OpenACCDirectiveKind::Update:
+ return true;
+ default:
+ return false;
+ }
+ }
+ case OpenACCClauseKind::Host: {
+ switch (DirectiveKind) {
+ case OpenACCDirectiveKind::Update:
+ return true;
+ default:
+ return false;
+ }
+ }
}
default:
@@ -1040,6 +1056,28 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitPresentClause(
Clause.getVarList(), Clause.getEndLoc());
}
+OpenACCClause *SemaOpenACCClauseVisitor::VisitHostClause(
+ SemaOpenACC::OpenACCParsedClause &Clause) {
+ // 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 OpenACCHostClause::Create(Ctx, Clause.getBeginLoc(),
+ Clause.getLParenLoc(), Clause.getVarList(),
+ Clause.getEndLoc());
+}
+
+OpenACCClause *SemaOpenACCClauseVisitor::VisitDeviceClause(
+ SemaOpenACC::OpenACCParsedClause &Clause) {
+ // 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 OpenACCDeviceClause::Create(Ctx, Clause.getBeginLoc(),
+ Clause.getLParenLoc(), Clause.getVarList(),
+ Clause.getEndLoc());
+}
+
OpenACCClause *SemaOpenACCClauseVisitor::VisitCopyClause(
SemaOpenACC::OpenACCParsedClause &Clause) {
// Restrictions only properly implemented on 'compute'/'combined'/'data'
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index d00ad5a35e8235..4a3c739ecbeab8 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -11730,6 +11730,30 @@ void OpenACCClauseTransform<Derived>::VisitPrivateClause(
ParsedClause.getEndLoc());
}
+template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitHostClause(
+ const OpenACCHostClause &C) {
+ ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+ /*IsReadOnly=*/false, /*IsZero=*/false);
+
+ NewClause = OpenACCHostClause::Create(
+ Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
+ ParsedClause.getLParenLoc(), ParsedClause.getVarList(),
+ ParsedClause.getEndLoc());
+}
+
+template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitDeviceClause(
+ const OpenACCDeviceClause &C) {
+ ParsedClause.setVarListDetails(VisitVarList(C.getVarList()),
+ /*IsReadOnly=*/false, /*IsZero=*/false);
+
+ NewClause = OpenACCDeviceClause::Create(
+ Self.getSema().getASTContext(), ParsedClause.getBeginLoc(),
+ ParsedClause.getLParenLoc(), ParsedClause.getVarList(),
+ ParsedClause.getEndLoc());
+}
+
template <typename Derived>
void OpenACCClauseTransform<Derived>::VisitFirstPrivateClause(
const OpenACCFirstPrivateClause &C) {
diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp
index 0368990ca150df..b53f99732cacce 100644
--- a/clang/lib/Serialization/ASTReader.cpp
+++ b/clang/lib/Serialization/ASTReader.cpp
@@ -12439,6 +12439,18 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
return OpenACCPrivateClause::Create(getContext(), BeginLoc, LParenLoc,
VarList, EndLoc);
}
+ case OpenACCClauseKind::Host: {
+ SourceLocation LParenLoc = readSourceLocation();
+ llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
+ return OpenACCHostClause::Create(getContext(), BeginLoc, LParenLoc, VarList,
+ EndLoc);
+ }
+ case OpenACCClauseKind::Device: {
+ SourceLocation LParenLoc = readSourceLocation();
+ llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
+ return OpenACCDeviceClause::Create(getContext(), BeginLoc, LParenLoc,
+ VarList, EndLoc);
+ }
case OpenACCClauseKind::FirstPrivate: {
SourceLocation LParenLoc = readSourceLocation();
llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
@@ -12611,9 +12623,7 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
}
case OpenACCClauseKind::NoHost:
- case OpenACCClauseKind::Device:
case OpenACCClauseKind::DeviceResident:
- case OpenACCClauseKind::Host:
case OpenACCClauseKind::Link:
case OpenACCClauseKind::Bind:
case OpenACCClauseKind::Invalid:
diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp
index 8d9396e28ed50a..39004fd4d4c376 100644
--- a/clang/lib/Serialization/ASTWriter.cpp
+++ b/clang/lib/Serialization/ASTWriter.cpp
@@ -8371,6 +8371,18 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
writeOpenACCVarList(PC);
return;
}
+ case OpenACCClauseKind::Host: {
+ const auto *HC = cast<OpenACCHostClause>(C);
+ writeSourceLocation(HC->getLParenLoc());
+ writeOpenACCVarList(HC);
+ return;
+ }
+ case OpenACCClauseKind::Device: {
+ const auto *DC = cast<OpenACCDeviceClause>(C);
+ writeSourceLocation(DC->getLParenLoc());
+ writeOpenACCVarList(DC);
+ return;
+ }
case OpenACCClauseKind::FirstPrivate: {
const auto *FPC = cast<OpenACCFirstPrivateClause>(C);
writeSourceLocation(FPC->getLParenLoc());
@@ -8544,9 +8556,7 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
}
case OpenACCClauseKind::NoHost:
- case OpenACCClauseKind::Device:
case OpenACCClauseKind::DeviceResident:
- case OpenACCClauseKind::Host:
case OpenACCClauseKind::Link:
case OpenACCClauseKind::Bind:
case OpenACCClauseKind::Invalid:
diff --git a/clang/test/AST/ast-print-openacc-update-construct.cpp b/clang/test/AST/ast-print-openacc-update-construct.cpp
index a7f5b2a42285be..0d09c829929dc7 100644
--- a/clang/test/AST/ast-print-openacc-update-construct.cpp
+++ b/clang/test/AST/ast-print-openacc-update-construct.cpp
@@ -38,4 +38,10 @@ void uses(bool cond) {
// 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])
+
+// CHECK: #pragma acc update host(I, iPtr, array, array[1], array[1:2])
+#pragma acc update host (I, iPtr, array, array[1], array[1:2])
+
+// CHECK: #pragma acc update device(I, iPtr, array, array[1], array[1:2])
+#pragma acc update device(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 73a09697710f98..930cc1e4c0353a 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -538,22 +538,18 @@ void VarListClauses() {
#pragma acc serial link(s.array[s.value : 5], s.value), self
for(int i = 0; i < 5;++i) {}
- // expected-error at +2{{expected ','}}
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented, clause ignored}}
-#pragma acc serial host(s.array[s.value] s.array[s.value :5] ), self
+ // expected-error at +1{{expected ','}}
+#pragma acc update host(s.array[s.value] s.array[s.value :5] )
for(int i = 0; i < 5;++i) {}
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented, clause ignored}}
-#pragma acc serial host(s.array[s.value : 5], s.value), self
+#pragma acc update host(s.array[s.value : 5], s.value)
for(int i = 0; i < 5;++i) {}
- // expected-error at +2{{expected ','}}
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented, clause ignored}}
-#pragma acc serial device(s.array[s.value] s.array[s.value :5] ), self
+ // expected-error at +1{{expected ','}}
+#pragma acc update device(s.array[s.value] s.array[s.value :5] )
for(int i = 0; i < 5;++i) {}
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented, clause ignored}}
-#pragma acc serial device(s.array[s.value : 5], s.value), self
+#pragma acc update device(s.array[s.value : 5], s.value)
for(int i = 0; i < 5;++i) {}
// expected-error at +1{{expected ','}}
diff --git a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
index 9a98f5e1659b8d..9e74ce27ffbd98 100644
--- a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
+++ b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c
@@ -75,7 +75,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop auto detach(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop auto device(VarPtr)
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop auto deviceptr(VarPtr)
@@ -85,7 +85,7 @@ void uses() {
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop auto firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop auto host(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -192,7 +192,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop detach(Var) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop device(VarPtr) auto
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop deviceptr(VarPtr) auto
@@ -202,7 +202,7 @@ void uses() {
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop firstprivate(Var) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop host(Var) auto
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -310,7 +310,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop independent detach(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop independent device(VarPtr)
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop independent deviceptr(VarPtr)
@@ -320,7 +320,7 @@ void uses() {
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop independent firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop independent host(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -427,7 +427,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop detach(Var) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop device(VarPtr) independent
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop deviceptr(VarPtr) independent
@@ -437,7 +437,7 @@ void uses() {
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop firstprivate(Var) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop host(Var) independent
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -553,7 +553,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop seq detach(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop seq device(VarPtr)
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop seq deviceptr(VarPtr)
@@ -563,7 +563,7 @@ void uses() {
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop seq firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop seq host(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -676,7 +676,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop detach(Var) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop device(VarPtr) seq
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop deviceptr(VarPtr) seq
@@ -686,7 +686,7 @@ void uses() {
for(unsigned i = 0; i < 5; ++i);
#pragma acc parallel loop firstprivate(Var) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop host(Var) seq
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
index 8072ae7de4ffc0..c185f607548ff8 100644
--- a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c
@@ -100,8 +100,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'kernels loop' directive}}
#pragma acc kernels loop device_type(*) detach(Var)
for(int i = 0; i < 5; ++i);
- // expected-error at +2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'parallel loop' construct}}
- // expected-note at +1{{previous clause is here}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'parallel loop' directive}}
#pragma acc parallel loop device_type(*) device(VarPtr)
for(int i = 0; i < 5; ++i);
// expected-error at +2{{OpenACC clause 'deviceptr' may not follow a 'device_type' clause in a 'serial loop' construct}}
@@ -116,8 +115,7 @@ void uses() {
// expected-note at +1{{previous clause is here}}
#pragma acc parallel loop device_type(*) firstprivate(Var)
for(int i = 0; i < 5; ++i);
- // expected-error at +2{{OpenACC clause 'host' may not follow a 'device_type' clause in a 'serial loop' construct}}
- // expected-note at +1{{previous clause is here}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'serial loop' directive}}
#pragma acc serial loop device_type(*) host(Var)
for(int i = 0; i < 5; ++i);
// expected-error at +2{{OpenACC clause 'link' may not follow a 'device_type' clause in a 'parallel loop' construct}}
diff --git a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
index 33f433ce4c519f..4290fb76656859 100644
--- a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c
@@ -106,8 +106,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'kernels' directive}}
#pragma acc kernels device_type(*) detach(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'kernels' construct}}
- // expected-note at +1{{previous clause is here}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'kernels' directive}}
#pragma acc kernels device_type(*) device(VarPtr)
while(1);
// expected-error at +2{{OpenACC clause 'deviceptr' may not follow a 'device_type' clause in a 'kernels' construct}}
@@ -122,8 +121,7 @@ void uses() {
// expected-note at +1{{previous clause is here}}
#pragma acc parallel device_type(*) firstprivate(Var)
while(1);
- // expected-error at +2{{OpenACC clause 'host' may not follow a 'device_type' clause in a 'kernels' construct}}
- // expected-note at +1{{previous clause is here}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'kernels' directive}}
#pragma acc kernels device_type(*) host(Var)
while(1);
// expected-error at +2{{OpenACC clause 'link' may not follow a 'device_type' clause in a 'kernels' construct}}
diff --git a/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c b/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c
index d9a4c61a81c7d7..f56a1267fbad10 100644
--- a/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c
+++ b/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c
@@ -80,7 +80,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'loop' directive}}
#pragma acc loop auto detach(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'loop' directive}}
#pragma acc loop auto device(VarPtr)
for(unsigned i = 0; i < 5; ++i);
// expected-error at +1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
@@ -92,7 +92,7 @@ void uses() {
// expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
#pragma acc loop auto firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'loop' directive}}
#pragma acc loop auto host(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -214,7 +214,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'loop' directive}}
#pragma acc loop detach(Var) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'loop' directive}}
#pragma acc loop device(VarPtr) auto
for(unsigned i = 0; i < 5; ++i);
// expected-error at +1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
@@ -226,7 +226,7 @@ void uses() {
// expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
#pragma acc loop firstprivate(Var) auto
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'loop' directive}}
#pragma acc loop host(Var) auto
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -349,7 +349,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'loop' directive}}
#pragma acc loop independent detach(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'loop' directive}}
#pragma acc loop independent device(VarPtr)
for(unsigned i = 0; i < 5; ++i);
// expected-error at +1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
@@ -361,7 +361,7 @@ void uses() {
// expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
#pragma acc loop independent firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'loop' directive}}
#pragma acc loop independent host(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -483,7 +483,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'loop' directive}}
#pragma acc loop detach(Var) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'loop' directive}}
#pragma acc loop device(VarPtr) independent
for(unsigned i = 0; i < 5; ++i);
// expected-error at +1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
@@ -495,7 +495,7 @@ void uses() {
// expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
#pragma acc loop firstprivate(Var) independent
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'loop' directive}}
#pragma acc loop host(Var) independent
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -626,7 +626,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'loop' directive}}
#pragma acc loop seq detach(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'loop' directive}}
#pragma acc loop seq device(VarPtr)
for(unsigned i = 0; i < 5; ++i);
// expected-error at +1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
@@ -638,7 +638,7 @@ void uses() {
// expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
#pragma acc loop seq firstprivate(Var)
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'loop' directive}}
#pragma acc loop seq host(Var)
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
@@ -766,7 +766,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'loop' directive}}
#pragma acc loop detach(Var) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'loop' directive}}
#pragma acc loop device(VarPtr) seq
for(unsigned i = 0; i < 5; ++i);
// expected-error at +1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
@@ -778,7 +778,7 @@ void uses() {
// expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
#pragma acc loop firstprivate(Var) seq
for(unsigned i = 0; i < 5; ++i);
- // expected-warning at +1{{OpenACC clause 'host' not yet implemented}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'loop' directive}}
#pragma acc loop host(Var) seq
for(unsigned i = 0; i < 5; ++i);
// expected-warning at +1{{OpenACC clause 'link' not yet implemented}}
diff --git a/clang/test/SemaOpenACC/loop-construct-device_type-clause.c b/clang/test/SemaOpenACC/loop-construct-device_type-clause.c
index f16f17f5d68108..2c1189bc647d0a 100644
--- a/clang/test/SemaOpenACC/loop-construct-device_type-clause.c
+++ b/clang/test/SemaOpenACC/loop-construct-device_type-clause.c
@@ -92,8 +92,7 @@ void uses() {
// expected-error at +1{{OpenACC 'detach' clause is not valid on 'loop' directive}}
#pragma acc loop device_type(*) detach(Var)
for(int i = 0; i < 5; ++i);
- // expected-error at +2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'loop' construct}}
- // expected-note at +1{{previous clause is here}}
+ // expected-error at +1{{OpenACC 'device' clause is not valid on 'loop' directive}}
#pragma acc loop device_type(*) device(VarPtr)
for(int i = 0; i < 5; ++i);
// expected-error at +1{{OpenACC 'deviceptr' clause is not valid on 'loop' directive}}
@@ -106,8 +105,7 @@ void uses() {
// expected-error at +1{{OpenACC 'firstprivate' clause is not valid on 'loop' directive}}
#pragma acc loop device_type(*) firstprivate(Var)
for(int i = 0; i < 5; ++i);
- // expected-error at +2{{OpenACC clause 'host' may not follow a 'device_type' clause in a 'loop' construct}}
- // expected-note at +1{{previous clause is here}}
+ // expected-error at +1{{OpenACC 'host' clause is not valid on 'loop' directive}}
#pragma acc loop device_type(*) host(Var)
for(int i = 0; i < 5; ++i);
// expected-error at +2{{OpenACC clause 'link' may not follow a 'device_type' clause in a 'loop' construct}}
diff --git a/clang/test/SemaOpenACC/update-construct-ast.cpp b/clang/test/SemaOpenACC/update-construct-ast.cpp
index 9048e8823f5f5f..5582bde7a64c2d 100644
--- a/clang/test/SemaOpenACC/update-construct-ast.cpp
+++ b/clang/test/SemaOpenACC/update-construct-ast.cpp
@@ -89,6 +89,33 @@ void NormalFunc() {
// CHECK-NEXT: DeclRefExpr{{.*}}'GlobalArray' 'short[5]'
// CHECK-NEXT: IntegerLiteral{{.*}} 0
// CHECK-NEXT: IntegerLiteral{{.*}} 1
+
+#pragma acc update host(Global, GlobalArray, GlobalArray[0], GlobalArray[0:1]) device(Global, GlobalArray, GlobalArray[0], GlobalArray[0:1])
+ // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
+ // CHECK-NEXT: host 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
+ // CHECK-NEXT: device 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>
@@ -163,6 +190,29 @@ void TemplFunc(T t) {
// CHECK-NEXT: IntegerLiteral{{.*}}0
// CHECK-NEXT: IntegerLiteral{{.*}}1
+#pragma acc update host(Local, LocalArray, LocalArray[0], LocalArray[0:1]) device(Local, LocalArray, LocalArray[0], LocalArray[0:1])
+ // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
+ // CHECK-NEXT: host 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
+ // CHECK-NEXT: device 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'
@@ -255,6 +305,32 @@ void TemplFunc(T t) {
// CHECK-NEXT: DeclRefExpr{{.*}} 'LocalArray' 'decltype(SomeStruct::value)[5]'
// CHECK-NEXT: IntegerLiteral{{.*}}0
// CHECK-NEXT: IntegerLiteral{{.*}}1
+
+ // CHECK-NEXT: OpenACCUpdateConstruct{{.*}}update
+ // CHECK-NEXT: host 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{{.*}}0
+ // CHECK-NEXT: IntegerLiteral{{.*}}1
+ // CHECK-NEXT: device 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{{.*}}0
+ // CHECK-NEXT: IntegerLiteral{{.*}}1
}
struct SomeStruct{
diff --git a/clang/test/SemaOpenACC/update-construct.cpp b/clang/test/SemaOpenACC/update-construct.cpp
index 2abd7a30eda888..8b65da69e42b9f 100644
--- a/clang/test/SemaOpenACC/update-construct.cpp
+++ b/clang/test/SemaOpenACC/update-construct.cpp
@@ -10,9 +10,7 @@ void uses() {
#pragma acc update if(true) self(Var)
#pragma acc update if_present self(Var)
#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-error at +2{{OpenACC clause 'if' may not follow a 'device_type' clause in a 'update' construct}}
@@ -56,35 +54,29 @@ void uses() {
// Cannot be the body of an 'if', 'while', 'do', 'switch', or
// 'label'.
- // expected-error at +3{{OpenACC 'update' construct may not appear in place of the statement following an if statement}}
+ // expected-error at +2{{OpenACC 'update' construct may not appear in place of the statement following an if statement}}
if (true)
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
#pragma acc update device(Var)
- // expected-error at +3{{OpenACC 'update' construct may not appear in place of the statement following a while statement}}
+ // expected-error at +2{{OpenACC 'update' construct may not appear in place of the statement following a while statement}}
while (true)
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
#pragma acc update device(Var)
- // expected-error at +3{{OpenACC 'update' construct may not appear in place of the statement following a do statement}}
+ // expected-error at +2{{OpenACC 'update' construct may not appear in place of the statement following a do statement}}
do
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
#pragma acc update device(Var)
while (true);
- // expected-error at +3{{OpenACC 'update' construct may not appear in place of the statement following a switch statement}}
+ // expected-error at +2{{OpenACC 'update' construct may not appear in place of the statement following a switch statement}}
switch(Var)
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
#pragma acc update device(Var)
- // expected-error at +3{{OpenACC 'update' construct may not appear in place of the statement following a label statement}}
+ // expected-error at +2{{OpenACC 'update' construct may not appear in place of the statement following a label statement}}
LABEL:
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
#pragma acc update device(Var)
// For loops are OK.
for (;;)
- // expected-warning at +1{{OpenACC clause 'device' not yet implemented}}
#pragma acc update device(Var)
// Checking for 'async', which requires an 'int' expression.
@@ -132,11 +124,19 @@ void varlist_restrictions_templ() {
// 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 host(iArray[0:1])
+#pragma acc update device(iArray[0:1])
#pragma acc update self(Array[0:1])
+#pragma acc update host(Array[0:1])
+#pragma acc update device(Array[0:1])
// expected-error at +1{{OpenACC sub-array is not allowed here}}
#pragma acc update self(Array[0:1].MemberOfComp)
+ // expected-error at +1{{OpenACC sub-array is not allowed here}}
+#pragma acc update host(Array[0:1].MemberOfComp)
+ // expected-error at +1{{OpenACC sub-array is not allowed here}}
+#pragma acc update device(Array[0:1].MemberOfComp)
}
void varlist_restrictions() {
@@ -149,19 +149,33 @@ void varlist_restrictions() {
int *LocalPtr;
#pragma acc update self(LocalInt, LocalPtr, Single)
+#pragma acc update host(LocalInt, LocalPtr, Single)
+#pragma acc update device(LocalInt, LocalPtr, Single)
#pragma acc update self(Single.MemberOfComp)
+#pragma acc update host(Single.MemberOfComp)
+#pragma acc update device(Single.MemberOfComp)
#pragma acc update self(Single.Array[0:1])
+#pragma acc update host(Single.Array[0:1])
+#pragma acc update device(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 host(iArray[0:1])
+#pragma acc update device(iArray[0:1])
#pragma acc update self(Array[0:1])
+#pragma acc update host(Array[0:1])
+#pragma acc update device(Array[0:1])
// expected-error at +1{{OpenACC sub-array is not allowed here}}
#pragma acc update self(Array[0:1].MemberOfComp)
+ // expected-error at +1{{OpenACC sub-array is not allowed here}}
+#pragma acc update host(Array[0:1].MemberOfComp)
+ // expected-error at +1{{OpenACC sub-array is not allowed here}}
+#pragma acc update device(Array[0:1].MemberOfComp)
}
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index 5e51fc4e2f66c2..e175aab4499fff 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2877,6 +2877,14 @@ void OpenACCClauseEnqueue::VisitPrivateClause(const OpenACCPrivateClause &C) {
VisitVarList(C);
}
+void OpenACCClauseEnqueue::VisitHostClause(const OpenACCHostClause &C) {
+ VisitVarList(C);
+}
+
+void OpenACCClauseEnqueue::VisitDeviceClause(const OpenACCDeviceClause &C) {
+ VisitVarList(C);
+}
+
void OpenACCClauseEnqueue::VisitFirstPrivateClause(
const OpenACCFirstPrivateClause &C) {
VisitVarList(C);
More information about the cfe-commits
mailing list