[clang] 054f7c0 - [OpenACC] Implement copy clause for compute constructs.
via cfe-commits
cfe-commits at lists.llvm.org
Fri May 3 07:20:47 PDT 2024
Author: erichkeane
Date: 2024-05-03T07:20:41-07:00
New Revision: 054f7c0565410f246f3e003ec18684bd526de64c
URL: https://github.com/llvm/llvm-project/commit/054f7c0565410f246f3e003ec18684bd526de64c
DIFF: https://github.com/llvm/llvm-project/commit/054f7c0565410f246f3e003ec18684bd526de64c.diff
LOG: [OpenACC] Implement copy clause for compute constructs.
Like present, no_create, and first_private, copy is a clause that takes
just a var-list, and follows the same rules as the others.
The one unique part of this clause is that it ALSO supports two
deprecated/backwards-compatibility spellings, so this patch adds them
and implements them.
Added:
clang/test/SemaOpenACC/compute-construct-copy-clause.c
clang/test/SemaOpenACC/compute-construct-copy-clause.cpp
Modified:
clang/include/clang/AST/OpenACCClause.h
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/include/clang/Basic/OpenACCClauses.def
clang/include/clang/Basic/OpenACCKinds.h
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/ParserOpenACC/parse-clauses.c
clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
clang/tools/libclang/CIndex.cpp
Removed:
################################################################################
diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h
index 369129a35d8128..ba20b2c811a8d8 100644
--- a/clang/include/clang/AST/OpenACCClause.h
+++ b/clang/include/clang/AST/OpenACCClause.h
@@ -351,6 +351,30 @@ class OpenACCPresentClause final
ArrayRef<Expr *> VarList, SourceLocation EndLoc);
};
+class OpenACCCopyClause final
+ : public OpenACCClauseWithVarList,
+ public llvm::TrailingObjects<OpenACCCopyClause, Expr *> {
+
+ OpenACCCopyClause(OpenACCClauseKind Spelling, SourceLocation BeginLoc,
+ SourceLocation LParenLoc, ArrayRef<Expr *> VarList,
+ SourceLocation EndLoc)
+ : OpenACCClauseWithVarList(Spelling, BeginLoc, LParenLoc, EndLoc) {
+ assert((Spelling == OpenACCClauseKind::Copy ||
+ Spelling == OpenACCClauseKind::PCopy ||
+ Spelling == OpenACCClauseKind::PresentOrCopy) &&
+ "Invalid clause kind for copy-clause");
+ std::uninitialized_copy(VarList.begin(), VarList.end(),
+ getTrailingObjects<Expr *>());
+ setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size()));
+ }
+
+public:
+ static OpenACCCopyClause *
+ Create(const ASTContext &C, OpenACCClauseKind Spelling,
+ SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc);
+};
+
template <class Impl> class OpenACCClauseVisitor {
Impl &getDerived() { return static_cast<Impl &>(*this); }
@@ -369,6 +393,10 @@ template <class Impl> class OpenACCClauseVisitor {
case OpenACCClauseKind::CLAUSE_NAME: \
Visit##CLAUSE_NAME##Clause(*cast<OpenACC##CLAUSE_NAME##Clause>(C)); \
return;
+#define CLAUSE_ALIAS(ALIAS_NAME, CLAUSE_NAME) \
+ case OpenACCClauseKind::ALIAS_NAME: \
+ Visit##CLAUSE_NAME##Clause(*cast<OpenACC##CLAUSE_NAME##Clause>(C)); \
+ return;
#include "clang/Basic/OpenACCClauses.def"
default:
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index b5ba4bc889f2e3..6c4d92790afc51 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12333,4 +12333,8 @@ def err_acc_subarray_out_of_range
def err_acc_subarray_base_plus_length_out_of_range
: Error<"OpenACC sub-array specified range [%0:%1] would be out of the "
"range of the subscripted array size of %2">;
+def warn_acc_deprecated_alias_name
+ : Warning<"OpenACC clause name '%0' is a deprecated clause name and is "
+ "now an alias for '%1'">,
+ InGroup<DiagGroup<"openacc-deprecated-clause-alias">>;
} // end of sema component.
diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def
index d5f14d8646a4be..5cb421975c95a5 100644
--- a/clang/include/clang/Basic/OpenACCClauses.def
+++ b/clang/include/clang/Basic/OpenACCClauses.def
@@ -14,7 +14,16 @@
// as used in Clang source (so `Default` instead of `default`).
//
// VISIT_CLAUSE(CLAUSE_NAME)
+//
+// CLAUSE_ALIAS(ALIAS_NAME, CLAUSE_NAME)
+
+#ifndef CLAUSE_ALIAS
+#define CLAUSE_ALIAS(ALIAS_NAME, CLAUSE_NAME)
+#endif
+VISIT_CLAUSE(Copy)
+CLAUSE_ALIAS(PCopy, Copy)
+CLAUSE_ALIAS(PresentOrCopy, Copy)
VISIT_CLAUSE(Default)
VISIT_CLAUSE(FirstPrivate)
VISIT_CLAUSE(If)
@@ -27,3 +36,4 @@ VISIT_CLAUSE(Self)
VISIT_CLAUSE(VectorLength)
#undef VISIT_CLAUSE
+#undef CLAUSE_ALIAS
diff --git a/clang/include/clang/Basic/OpenACCKinds.h b/clang/include/clang/Basic/OpenACCKinds.h
index e3f74178433285..e39f9e876d549c 100644
--- a/clang/include/clang/Basic/OpenACCKinds.h
+++ b/clang/include/clang/Basic/OpenACCKinds.h
@@ -189,6 +189,10 @@ enum class OpenACCClauseKind {
/// 'copy' clause, allowed on Compute and Combined Constructs, plus 'data' and
/// 'declare'.
Copy,
+ /// 'copy' clause alias 'pcopy'. Preserved for diagnostic purposes.
+ PCopy,
+ /// 'copy' clause alias 'present_or_copy'. Preserved for diagnostic purposes.
+ PresentOrCopy,
/// 'use_device' clause, allowed on 'host_data' construct.
UseDevice,
/// 'attach' clause, allowed on Compute and Combined constructs, plus 'data'
@@ -310,6 +314,12 @@ inline StreamTy &printOpenACCClauseKind(StreamTy &Out, OpenACCClauseKind K) {
case OpenACCClauseKind::Copy:
return Out << "copy";
+ case OpenACCClauseKind::PCopy:
+ return Out << "pcopy";
+
+ case OpenACCClauseKind::PresentOrCopy:
+ return Out << "present_or_copy";
+
case OpenACCClauseKind::UseDevice:
return Out << "use_device";
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index cd91847a9591e1..1c5127e79a6699 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -120,6 +120,9 @@ class SemaOpenACC : public SemaBase {
assert((ClauseKind == OpenACCClauseKind::Private ||
ClauseKind == OpenACCClauseKind::NoCreate ||
ClauseKind == OpenACCClauseKind::Present ||
+ ClauseKind == OpenACCClauseKind::Copy ||
+ ClauseKind == OpenACCClauseKind::PCopy ||
+ ClauseKind == OpenACCClauseKind::PresentOrCopy ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
"Parsed clause kind does not have a var-list");
return std::get<VarListDetails>(Details).VarList;
@@ -171,6 +174,9 @@ class SemaOpenACC : public SemaBase {
assert((ClauseKind == OpenACCClauseKind::Private ||
ClauseKind == OpenACCClauseKind::NoCreate ||
ClauseKind == OpenACCClauseKind::Present ||
+ ClauseKind == OpenACCClauseKind::Copy ||
+ ClauseKind == OpenACCClauseKind::PCopy ||
+ ClauseKind == OpenACCClauseKind::PresentOrCopy ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
"Parsed clause kind does not have a var-list");
Details = VarListDetails{{VarList.begin(), VarList.end()}};
@@ -180,6 +186,9 @@ class SemaOpenACC : public SemaBase {
assert((ClauseKind == OpenACCClauseKind::Private ||
ClauseKind == OpenACCClauseKind::NoCreate ||
ClauseKind == OpenACCClauseKind::Present ||
+ ClauseKind == OpenACCClauseKind::Copy ||
+ ClauseKind == OpenACCClauseKind::PCopy ||
+ ClauseKind == OpenACCClauseKind::PresentOrCopy ||
ClauseKind == OpenACCClauseKind::FirstPrivate) &&
"Parsed clause kind does not have a var-list");
Details = VarListDetails{std::move(VarList)};
diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp
index 88fcb9a84f1f0b..2a12a74899cdd5 100644
--- a/clang/lib/AST/OpenACCClause.cpp
+++ b/clang/lib/AST/OpenACCClause.cpp
@@ -76,6 +76,9 @@ OpenACCClause::child_range OpenACCClause::children() {
#define VISIT_CLAUSE(CLAUSE_NAME) \
case OpenACCClauseKind::CLAUSE_NAME: \
return cast<OpenACC##CLAUSE_NAME##Clause>(this)->children();
+#define CLAUSE_ALIAS(ALIAS_NAME, CLAUSE_NAME) \
+ case OpenACCClauseKind::ALIAS_NAME: \
+ return cast<OpenACC##CLAUSE_NAME##Clause>(this)->children();
#include "clang/Basic/OpenACCClauses.def"
}
@@ -173,6 +176,16 @@ OpenACCPresentClause *OpenACCPresentClause::Create(const ASTContext &C,
return new (Mem) OpenACCPresentClause(BeginLoc, LParenLoc, VarList, EndLoc);
}
+OpenACCCopyClause *
+OpenACCCopyClause::Create(const ASTContext &C, OpenACCClauseKind Spelling,
+ SourceLocation BeginLoc, SourceLocation LParenLoc,
+ ArrayRef<Expr *> VarList, SourceLocation EndLoc) {
+ void *Mem =
+ C.Allocate(OpenACCCopyClause::totalSizeToAlloc<Expr *>(VarList.size()));
+ return new (Mem)
+ OpenACCCopyClause(Spelling, BeginLoc, LParenLoc, VarList, EndLoc);
+}
+
//===----------------------------------------------------------------------===//
// OpenACC clauses printing methods
//===----------------------------------------------------------------------===//
@@ -249,3 +262,10 @@ void OpenACCClausePrinter::VisitPresentClause(const OpenACCPresentClause &C) {
[&](const Expr *E) { printExpr(E); });
OS << ")";
}
+
+void OpenACCClausePrinter::VisitCopyClause(const OpenACCCopyClause &C) {
+ OS << C.getClauseKind() << '(';
+ 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 b4ca81f5e25f16..6fe4f673a59d11 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -2492,6 +2492,11 @@ void OpenACCClauseProfiler::VisitIfClause(const OpenACCIfClause &Clause) {
Profiler.VisitStmt(Clause.getConditionExpr());
}
+void OpenACCClauseProfiler::VisitCopyClause(const OpenACCCopyClause &Clause) {
+ for (auto *E : Clause.getVarList())
+ Profiler.VisitStmt(E);
+}
+
void OpenACCClauseProfiler::VisitSelfClause(const OpenACCSelfClause &Clause) {
if (Clause.hasConditionExpr())
Profiler.VisitStmt(Clause.getConditionExpr());
diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp
index 7d3ea76fbd61c6..d383d9fe170423 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -397,6 +397,9 @@ void TextNodeDumper::Visit(const OpenACCClause *C) {
case OpenACCClauseKind::Default:
OS << '(' << cast<OpenACCDefaultClause>(C)->getDefaultClauseKind() << ')';
break;
+ case OpenACCClauseKind::Copy:
+ case OpenACCClauseKind::PCopy:
+ case OpenACCClauseKind::PresentOrCopy:
case OpenACCClauseKind::If:
case OpenACCClauseKind::FirstPrivate:
case OpenACCClauseKind::NoCreate:
diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp
index a1074abd82faa3..bc785fa7136aaa 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -102,6 +102,8 @@ OpenACCClauseKind getOpenACCClauseKind(Token Tok) {
.Case("create", OpenACCClauseKind::Create)
.Case("collapse", OpenACCClauseKind::Collapse)
.Case("copy", OpenACCClauseKind::Copy)
+ .Case("pcopy", OpenACCClauseKind::PCopy)
+ .Case("present_or_copy", OpenACCClauseKind::PresentOrCopy)
.Case("copyin", OpenACCClauseKind::CopyIn)
.Case("copyout", OpenACCClauseKind::CopyOut)
.Case("default", OpenACCClauseKind::Default)
@@ -489,6 +491,8 @@ ClauseParensKind getClauseParensKind(OpenACCDirectiveKind DirKind,
case OpenACCClauseKind::If:
case OpenACCClauseKind::Create:
case OpenACCClauseKind::Copy:
+ case OpenACCClauseKind::PCopy:
+ case OpenACCClauseKind::PresentOrCopy:
case OpenACCClauseKind::CopyIn:
case OpenACCClauseKind::CopyOut:
case OpenACCClauseKind::UseDevice:
@@ -920,7 +924,6 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
assert(DirKind == OpenACCDirectiveKind::Update);
[[fallthrough]];
case OpenACCClauseKind::Attach:
- case OpenACCClauseKind::Copy:
case OpenACCClauseKind::Delete:
case OpenACCClauseKind::Detach:
case OpenACCClauseKind::Device:
@@ -931,6 +934,9 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
case OpenACCClauseKind::UseDevice:
ParseOpenACCVarList();
break;
+ case OpenACCClauseKind::Copy:
+ case OpenACCClauseKind::PCopy:
+ case OpenACCClauseKind::PresentOrCopy:
case OpenACCClauseKind::FirstPrivate:
case OpenACCClauseKind::NoCreate:
case OpenACCClauseKind::Present:
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 3572d8f1090016..ee73447067b1ab 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -153,6 +153,23 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind,
default:
return false;
}
+
+ case OpenACCClauseKind::Copy:
+ case OpenACCClauseKind::PCopy:
+ case OpenACCClauseKind::PresentOrCopy:
+ switch (DirectiveKind) {
+ case OpenACCDirectiveKind::Parallel:
+ case OpenACCDirectiveKind::Serial:
+ case OpenACCDirectiveKind::Kernels:
+ case OpenACCDirectiveKind::Data:
+ case OpenACCDirectiveKind::Declare:
+ 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;
@@ -413,6 +430,26 @@ SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses,
getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(),
Clause.getVarList(), Clause.getEndLoc());
}
+ case OpenACCClauseKind::PresentOrCopy:
+ case OpenACCClauseKind::PCopy:
+ Diag(Clause.getBeginLoc(), diag::warn_acc_deprecated_alias_name)
+ << Clause.getClauseKind() << OpenACCClauseKind::Copy;
+ LLVM_FALLTHROUGH;
+ case OpenACCClauseKind::Copy: {
+ // 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 OpenACCCopyClause::Create(
+ getASTContext(), Clause.getClauseKind(), Clause.getBeginLoc(),
+ Clause.getLParenLoc(), Clause.getVarList(), Clause.getEndLoc());
+ }
default:
break;
}
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index 91945fcf2d7faf..810247f8700f68 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -11264,6 +11264,17 @@ void OpenACCClauseTransform<Derived>::VisitPresentClause(
ParsedClause.getEndLoc());
}
+template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitCopyClause(
+ const OpenACCCopyClause &C) {
+ ParsedClause.setVarListDetails(VisitVarList(C.getVarList()));
+
+ NewClause = OpenACCCopyClause::Create(
+ Self.getSema().getASTContext(), ParsedClause.getClauseKind(),
+ ParsedClause.getBeginLoc(), ParsedClause.getLParenLoc(),
+ ParsedClause.getVarList(), ParsedClause.getEndLoc());
+}
+
template <typename Derived>
void OpenACCClauseTransform<Derived>::VisitNumWorkersClause(
const OpenACCNumWorkersClause &C) {
diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp
index ca10bda18f1724..9f4397906b8ce7 100644
--- a/clang/lib/Serialization/ASTReader.cpp
+++ b/clang/lib/Serialization/ASTReader.cpp
@@ -11853,6 +11853,14 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
return OpenACCPresentClause::Create(getContext(), BeginLoc, LParenLoc,
VarList, EndLoc);
}
+ case OpenACCClauseKind::PCopy:
+ case OpenACCClauseKind::PresentOrCopy:
+ case OpenACCClauseKind::Copy: {
+ SourceLocation LParenLoc = readSourceLocation();
+ llvm::SmallVector<Expr *> VarList = readOpenACCVarList();
+ return OpenACCCopyClause::Create(getContext(), ClauseKind, BeginLoc,
+ LParenLoc, VarList, EndLoc);
+ }
case OpenACCClauseKind::Finalize:
case OpenACCClauseKind::IfPresent:
case OpenACCClauseKind::Seq:
@@ -11861,7 +11869,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
case OpenACCClauseKind::Worker:
case OpenACCClauseKind::Vector:
case OpenACCClauseKind::NoHost:
- case OpenACCClauseKind::Copy:
case OpenACCClauseKind::UseDevice:
case OpenACCClauseKind::Attach:
case OpenACCClauseKind::Delete:
diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp
index e9698c28dcc9f2..118e4395d971af 100644
--- a/clang/lib/Serialization/ASTWriter.cpp
+++ b/clang/lib/Serialization/ASTWriter.cpp
@@ -7805,6 +7805,15 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
writeOpenACCVarList(PC);
return;
}
+ case OpenACCClauseKind::Copy:
+ case OpenACCClauseKind::PCopy:
+ case OpenACCClauseKind::PresentOrCopy: {
+ const auto *CC = cast<OpenACCCopyClause>(C);
+ writeSourceLocation(CC->getLParenLoc());
+ writeOpenACCVarList(CC);
+ return;
+ }
+
case OpenACCClauseKind::Finalize:
case OpenACCClauseKind::IfPresent:
case OpenACCClauseKind::Seq:
@@ -7813,7 +7822,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
case OpenACCClauseKind::Worker:
case OpenACCClauseKind::Vector:
case OpenACCClauseKind::NoHost:
- case OpenACCClauseKind::Copy:
case OpenACCClauseKind::UseDevice:
case OpenACCClauseKind::Attach:
case OpenACCClauseKind::Delete:
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index 9366fc06d112ae..c1287c329bc86d 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -453,13 +453,11 @@ void VarListClauses() {
#pragma acc serial copy(, seq
for(;;){}
- // expected-error at +2{{expected expression}}
- // expected-warning at +1{{OpenACC clause 'copy' not yet implemented, clause ignored}}
+ // expected-error at +1{{expected expression}}
#pragma acc serial copy()
for(;;){}
- // expected-error at +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
+ // expected-error at +2{{expected expression}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(), seq
for(;;){}
@@ -467,62 +465,60 @@ void VarListClauses() {
struct Members s;
struct HasMembersArray HasMem;
- // 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(s.array[s.value]), seq
for(;;){}
- // 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(s.array[s.value], s.array[s.value :5] ), seq
for(;;){}
- // 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].array[1]), seq
for(;;){}
- // 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].array[1:4]), seq
for(;;){}
- // expected-error at +3{{OpenACC sub-array is not allowed here}}
- // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
+ // expected-error at +2{{OpenACC sub-array is not allowed here}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[1:3].array[1]), seq
for(;;){}
- // expected-error at +3{{OpenACC sub-array is not allowed here}}
- // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
+ // expected-error at +2{{OpenACC sub-array is not allowed here}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[1:3].array[1:2]), seq
for(;;){}
- // 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 +3{{expected expression}}
- // expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
+ // expected-error at +2{{expected expression}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[::]), seq
for(;;){}
- // 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-error at +4{{expected expression}}
+ // expected-error at +3{{expected ']'}}
+ // expected-note at +2{{to match this '['}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[: :]), seq
for(;;){}
- // 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(;;){}
+ // expected-warning at +1{{OpenACC clause name 'pcopy' is a deprecated clause name and is now an alias for 'copy'}}
+#pragma acc serial pcopy(HasMem.MemArr[3:])
+ for(;;){}
+
+ // expected-warning at +1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}}
+#pragma acc serial present_or_copy(HasMem.MemArr[3:])
+ for(;;){}
+
// expected-error at +3{{expected ','}}
// expected-warning at +2{{OpenACC clause 'use_device' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
diff --git a/clang/test/SemaOpenACC/compute-construct-copy-clause.c b/clang/test/SemaOpenACC/compute-construct-copy-clause.c
new file mode 100644
index 00000000000000..accbe43cea406f
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-copy-clause.c
@@ -0,0 +1,62 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ void *PointerMember;
+} Complete;
+void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+ int LocalInt;
+ short *LocalPointer;
+ float LocalArray[5];
+ Complete LocalComposite;
+ // Check Appertainment:
+#pragma acc parallel copy(LocalInt)
+ while(1);
+#pragma acc serial copy(LocalInt)
+ while(1);
+#pragma acc kernels copy(LocalInt)
+ while(1);
+
+ // expected-warning at +1{{OpenACC clause name 'pcopy' is a deprecated clause name and is now an alias for 'copy'}}
+#pragma acc parallel pcopy(LocalInt)
+ while(1);
+
+ // expected-warning at +1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}}
+#pragma acc parallel present_or_copy(LocalInt)
+ while(1);
+
+ // Valid cases:
+#pragma acc parallel copy(LocalInt, LocalPointer, LocalArray)
+ while(1);
+#pragma acc parallel copy(LocalArray[2:1])
+ while(1);
+
+#pragma acc parallel copy(LocalComposite.ScalarMember, LocalComposite.ScalarMember)
+ 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 copy(1 + 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 copy(+IntParam)
+ while(1);
+
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copy(PointerParam[2:])
+ while(1);
+
+ // expected-error at +1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel copy(ArrayParam[2:5])
+ while(1);
+
+ // expected-error at +2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copy((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 copy((float)ArrayParam[2])
+ while(1);
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-copy-clause.cpp b/clang/test/SemaOpenACC/compute-construct-copy-clause.cpp
new file mode 100644
index 00000000000000..16e78a43026a9f
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-copy-clause.cpp
@@ -0,0 +1,112 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+enum SomeE{};
+typedef struct IsComplete {
+ struct S { int A; } CompositeMember;
+ int ScalarMember;
+ float ArrayMember[5];
+ SomeE EnumMember;
+ char *PointerMember;
+} Complete;
+
+void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) {
+ int LocalInt;
+ char *LocalPointer;
+ float LocalArray[5];
+ // Check Appertainment:
+#pragma acc parallel copy(LocalInt)
+ while(1);
+#pragma acc serial copy(LocalInt)
+ while(1);
+#pragma acc kernels copy(LocalInt)
+ while(1);
+
+ // Valid cases:
+#pragma acc parallel copy(LocalInt, LocalPointer, LocalArray)
+ while(1);
+#pragma acc parallel copy(LocalArray[2:1])
+ while(1);
+
+ Complete LocalComposite2;
+#pragma acc parallel copy(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember)
+ 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 copy(1 + 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 copy(+IntParam)
+ while(1);
+
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copy(PointerParam[2:])
+ while(1);
+
+ // expected-error at +1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel copy(ArrayParam[2:5])
+ while(1);
+
+ // expected-error at +2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
+#pragma acc parallel copy((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 copy((float)ArrayParam[2])
+ while(1);
+}
+
+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 copy(+t)
+ 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 copy(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 copy(t, I)
+ while(true);
+
+#pragma acc parallel copy(arrayT)
+ while(true);
+
+#pragma acc parallel copy(TemplComp)
+ while(true);
+
+#pragma acc parallel copy(TemplComp.PointerMember[5])
+ while(true);
+ int *Pointer;
+#pragma acc parallel copy(Pointer[:I])
+ while(true);
+#pragma acc parallel copy(Pointer[:t])
+ while(true);
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel copy(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 copy(I)
+ while(true);
+
+#pragma acc parallel copy(NTTP_REF)
+ 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
index 7bdbaad977559a..d3de9f146d8db4 100644
--- a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
@@ -1,4 +1,4 @@
-// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
+// RUN: %clang_cc1 %s -fopenacc -Wno-openacc-deprecated-clause-alias -ast-dump | FileCheck %s
int Global;
short GlobalArray[5];
@@ -92,6 +92,23 @@ void NormalUses(float *PointerParam) {
// CHECK-NEXT: CXXBoolLiteralExpr
// CHECK-NEXT: NullStmt
+#pragma acc parallel copy(GlobalArray) pcopy(PointerParam[Global]) present_or_copy(Global)
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: copy clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
+ // CHECK-NEXT: pcopy 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: present_or_copy clause
+ // 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
@@ -243,6 +260,23 @@ void TemplUses(T t, U u, T*PointerParam) {
// CHECK-NEXT: CXXBoolLiteralExpr
// CHECK-NEXT: NullStmt
+#pragma acc parallel copy(t) pcopy(NTTP, u) present_or_copy(u[0:t])
+ while(true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: copy clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: pcopy clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &'
+ // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: present_or_copy 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
+
#pragma acc parallel private(u[0])
while(true);
// CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
@@ -364,6 +398,26 @@ void TemplUses(T t, U u, T*PointerParam) {
// CHECK-NEXT: CXXBoolLiteralExpr
// CHECK-NEXT: NullStmt
+//#pragma acc parallel copy(t) pcopy(NTTP, u) copy_or_present(u[0:t])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: copy clause
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: pcopy 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: present_or_copy clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
// #pragma acc parallel private(u[0])
// CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
// CHECK-NEXT: private clause
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index 76249e10a0a819..c02545f76ed2ec 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2828,6 +2828,9 @@ void OpenACCClauseEnqueue::VisitPresentClause(const OpenACCPresentClause &C) {
void OpenACCClauseEnqueue::VisitNoCreateClause(const OpenACCNoCreateClause &C) {
VisitVarList(C);
}
+void OpenACCClauseEnqueue::VisitCopyClause(const OpenACCCopyClause &C) {
+ VisitVarList(C);
+}
} // namespace
void EnqueueVisitor::EnqueueChildren(const OpenACCClause *C) {
More information about the cfe-commits
mailing list