Author: erichkeane Date: 2024-12-16T09:35:57-08:00 New Revision: fbb14dd97702db242a31e1b36ca8a3554a73c212
URL: https://github.com/llvm/llvm-project/commit/fbb14dd97702db242a31e1b36ca8a3554a73c212 DIFF: https://github.com/llvm/llvm-project/commit/fbb14dd97702db242a31e1b36ca8a3554a73c212.diff LOG: [OpenACC] Implement 'use_device' clause AST/Sema This is a clause that is only valid on 'host_data' constructs, and identifies variables which it should use the current device address. >From a Sema perspective, the only thing novel here is mild changes to how ActOnVar works for this clause, else this is very much like the rest of the 'var-list' clauses. Added: clang/test/SemaOpenACC/data-construct-use_device-ast.cpp clang/test/SemaOpenACC/data-construct-use_device-clause.c Modified: clang/include/clang/AST/OpenACCClause.h clang/include/clang/Basic/DiagnosticSemaKinds.td 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-data-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/data-construct-ast.cpp clang/test/SemaOpenACC/data-construct-async-clause.c clang/test/SemaOpenACC/data-construct-attach-clause.c clang/test/SemaOpenACC/data-construct-detach-clause.c clang/test/SemaOpenACC/data-construct-deviceptr-clause.c clang/test/SemaOpenACC/data-construct-finalize-clause.c clang/test/SemaOpenACC/data-construct-if-ast.cpp clang/test/SemaOpenACC/data-construct-if-clause.c clang/test/SemaOpenACC/data-construct-if_present-ast.cpp clang/test/SemaOpenACC/data-construct-if_present-clause.c clang/test/SemaOpenACC/data-construct-present-clause.c clang/test/SemaOpenACC/data-construct-wait-clause.c clang/test/SemaOpenACC/data-construct.cpp clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c clang/test/SemaOpenACC/loop-construct-device_type-clause.c clang/tools/libclang/CIndex.cpp Removed: ################################################################################ diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h index 93053c0e60758e..7a1b17cc4e44e3 100644 --- a/clang/include/clang/AST/OpenACCClause.h +++ b/clang/include/clang/AST/OpenACCClause.h @@ -788,6 +788,27 @@ class OpenACCDeleteClause final ArrayRef<Expr *> VarList, SourceLocation EndLoc); }; +class OpenACCUseDeviceClause final + : public OpenACCClauseWithVarList, + public llvm::TrailingObjects<OpenACCUseDeviceClause, Expr *> { + + OpenACCUseDeviceClause(SourceLocation BeginLoc, SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, SourceLocation EndLoc) + : OpenACCClauseWithVarList(OpenACCClauseKind::UseDevice, 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::UseDevice; + } + static OpenACCUseDeviceClause * + Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, SourceLocation EndLoc); +}; class OpenACCNoCreateClause final : public OpenACCClauseWithVarList, diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td index 811265151fa0da..77f84a89db2fc9 100644 --- a/clang/include/clang/Basic/DiagnosticSemaKinds.td +++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td @@ -12682,6 +12682,9 @@ def err_acc_not_a_var_ref : Error<"OpenACC variable is not a valid variable name, sub-array, array " "element,%select{| member of a composite variable,}0 or composite " "variable member">; +def err_acc_not_a_var_ref_use_device + : Error<"OpenACC variable in 'use_device' clause is not a valid variable " + "name or array name">; def err_acc_typecheck_subarray_value : Error<"OpenACC sub-array subscripted value is not an array or pointer">; def err_acc_subarray_function_type diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def index 600510e6980dae..c7ffac391e2026 100644 --- a/clang/include/clang/Basic/OpenACCClauses.def +++ b/clang/include/clang/Basic/OpenACCClauses.def @@ -58,6 +58,7 @@ VISIT_CLAUSE(Reduction) VISIT_CLAUSE(Self) VISIT_CLAUSE(Seq) VISIT_CLAUSE(Tile) +VISIT_CLAUSE(UseDevice) VISIT_CLAUSE(Vector) VISIT_CLAUSE(VectorLength) VISIT_CLAUSE(Wait) diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h index 58137d3a7e3f73..78056d03680017 100644 --- a/clang/include/clang/Sema/SemaOpenACC.h +++ b/clang/include/clang/Sema/SemaOpenACC.h @@ -400,6 +400,7 @@ class SemaOpenACC : public SemaBase { ClauseKind == OpenACCClauseKind::PresentOrCreate || ClauseKind == OpenACCClauseKind::Attach || ClauseKind == OpenACCClauseKind::Delete || + ClauseKind == OpenACCClauseKind::UseDevice || ClauseKind == OpenACCClauseKind::Detach || ClauseKind == OpenACCClauseKind::DevicePtr || ClauseKind == OpenACCClauseKind::Reduction || @@ -538,6 +539,7 @@ class SemaOpenACC : public SemaBase { ClauseKind == OpenACCClauseKind::PresentOrCreate || ClauseKind == OpenACCClauseKind::Attach || ClauseKind == OpenACCClauseKind::Delete || + ClauseKind == OpenACCClauseKind::UseDevice || ClauseKind == OpenACCClauseKind::Detach || ClauseKind == OpenACCClauseKind::DevicePtr || ClauseKind == OpenACCClauseKind::FirstPrivate) && @@ -576,6 +578,7 @@ class SemaOpenACC : public SemaBase { ClauseKind == OpenACCClauseKind::PresentOrCreate || ClauseKind == OpenACCClauseKind::Attach || ClauseKind == OpenACCClauseKind::Delete || + ClauseKind == OpenACCClauseKind::UseDevice || ClauseKind == OpenACCClauseKind::Detach || ClauseKind == OpenACCClauseKind::DevicePtr || ClauseKind == OpenACCClauseKind::FirstPrivate) && diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp index f836d30561e33b..fbc9f6d15fa7bf 100644 --- a/clang/lib/AST/OpenACCClause.cpp +++ b/clang/lib/AST/OpenACCClause.cpp @@ -33,6 +33,7 @@ bool OpenACCClauseWithVarList::classof(const OpenACCClause *C) { OpenACCFirstPrivateClause::classof(C) || OpenACCDevicePtrClause::classof(C) || OpenACCDeleteClause::classof(C) || + OpenACCUseDeviceClause::classof(C) || OpenACCDetachClause::classof(C) || OpenACCAttachClause::classof(C) || OpenACCNoCreateClause::classof(C) || OpenACCPresentClause::classof(C) || OpenACCCopyClause::classof(C) || @@ -298,6 +299,16 @@ OpenACCDeleteClause *OpenACCDeleteClause::Create(const ASTContext &C, return new (Mem) OpenACCDeleteClause(BeginLoc, LParenLoc, VarList, EndLoc); } +OpenACCUseDeviceClause *OpenACCUseDeviceClause::Create(const ASTContext &C, + SourceLocation BeginLoc, + SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, + SourceLocation EndLoc) { + void *Mem = C.Allocate( + OpenACCUseDeviceClause::totalSizeToAlloc<Expr *>(VarList.size())); + return new (Mem) OpenACCUseDeviceClause(BeginLoc, LParenLoc, VarList, EndLoc); +} + OpenACCDevicePtrClause *OpenACCDevicePtrClause::Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc, @@ -581,6 +592,14 @@ void OpenACCClausePrinter::VisitDeleteClause(const OpenACCDeleteClause &C) { OS << ")"; } +void OpenACCClausePrinter::VisitUseDeviceClause( + const OpenACCUseDeviceClause &C) { + OS << "use_device("; + llvm::interleaveComma(C.getVarList(), OS, + [&](const Expr *E) { printExpr(E); }); + OS << ")"; +} + void OpenACCClausePrinter::VisitDevicePtrClause( const OpenACCDevicePtrClause &C) { OS << "deviceptr("; diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp index 6160a69832e8aa..1fb238720ffb13 100644 --- a/clang/lib/AST/StmtProfile.cpp +++ b/clang/lib/AST/StmtProfile.cpp @@ -2515,6 +2515,11 @@ class OpenACCClauseProfiler } } + void VisitClauseWithVarList(const OpenACCClauseWithVarList &Clause) { + for (auto *E : Clause.getVarList()) + Profiler.VisitStmt(E); + } + #define VISIT_CLAUSE(CLAUSE_NAME) \ void Visit##CLAUSE_NAME##Clause(const OpenACC##CLAUSE_NAME##Clause &Clause); @@ -2532,25 +2537,21 @@ void OpenACCClauseProfiler::VisitIfClause(const OpenACCIfClause &Clause) { } void OpenACCClauseProfiler::VisitCopyClause(const OpenACCCopyClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitCopyInClause( const OpenACCCopyInClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitCopyOutClause( const OpenACCCopyOutClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitCreateClause( const OpenACCCreateClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitSelfClause(const OpenACCSelfClause &Clause) { @@ -2589,50 +2590,47 @@ void OpenACCClauseProfiler::VisitCollapseClause( void OpenACCClauseProfiler::VisitPrivateClause( const OpenACCPrivateClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitFirstPrivateClause( const OpenACCFirstPrivateClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitAttachClause( const OpenACCAttachClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitDetachClause( const OpenACCDetachClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitDeleteClause( const OpenACCDeleteClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitDevicePtrClause( const OpenACCDevicePtrClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitNoCreateClause( const OpenACCNoCreateClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitPresentClause( const OpenACCPresentClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); +} + +void OpenACCClauseProfiler::VisitUseDeviceClause( + const OpenACCUseDeviceClause &Clause) { + VisitClauseWithVarList(Clause); } void OpenACCClauseProfiler::VisitVectorLengthClause( @@ -2684,8 +2682,7 @@ void OpenACCClauseProfiler::VisitGangClause(const OpenACCGangClause &Clause) { void OpenACCClauseProfiler::VisitReductionClause( const OpenACCReductionClause &Clause) { - for (auto *E : Clause.getVarList()) - Profiler.VisitStmt(E); + VisitClauseWithVarList(Clause); } } // namespace diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp index 6040f34a4b9a5f..b5af10dd00b77c 100644 --- a/clang/lib/AST/TextNodeDumper.cpp +++ b/clang/lib/AST/TextNodeDumper.cpp @@ -425,6 +425,7 @@ void TextNodeDumper::Visit(const OpenACCClause *C) { case OpenACCClauseKind::Seq: case OpenACCClauseKind::Tile: case OpenACCClauseKind::Worker: + case OpenACCClauseKind::UseDevice: case OpenACCClauseKind::Vector: case OpenACCClauseKind::VectorLength: // The condition expression will be printed as a part of the 'children', diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp index 5130159f5d8ac8..570dba811aca86 100644 --- a/clang/lib/Parse/ParseOpenACC.cpp +++ b/clang/lib/Parse/ParseOpenACC.cpp @@ -1002,13 +1002,13 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams( case OpenACCClauseKind::DeviceResident: case OpenACCClauseKind::Host: case OpenACCClauseKind::Link: - case OpenACCClauseKind::UseDevice: ParseOpenACCVarList(ClauseKind); break; case OpenACCClauseKind::Attach: case OpenACCClauseKind::Delete: case OpenACCClauseKind::Detach: case OpenACCClauseKind::DevicePtr: + case OpenACCClauseKind::UseDevice: ParsedClause.setVarListDetails(ParseOpenACCVarList(ClauseKind), /*IsReadOnly=*/false, /*IsZero=*/false); break; diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp index 7af209aa155968..79822bf583195e 100644 --- a/clang/lib/Sema/SemaOpenACC.cpp +++ b/clang/lib/Sema/SemaOpenACC.cpp @@ -442,6 +442,15 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind, return false; } } + + case OpenACCClauseKind::UseDevice: { + switch (DirectiveKind) { + case OpenACCDirectiveKind::HostData: + return true; + default: + return false; + } + } } default: @@ -1085,6 +1094,14 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitDeleteClause( Clause.getEndLoc()); } +OpenACCClause *SemaOpenACCClauseVisitor::VisitUseDeviceClause( + SemaOpenACC::OpenACCParsedClause &Clause) { + // ActOnVar ensured that everything is a valid variable or array, so nothing + // left to do here. + return OpenACCUseDeviceClause::Create( + Ctx, Clause.getBeginLoc(), Clause.getLParenLoc(), Clause.getVarList(), + Clause.getEndLoc()); +} OpenACCClause *SemaOpenACCClauseVisitor::VisitDevicePtrClause( SemaOpenACC::OpenACCParsedClause &Clause) { @@ -2431,6 +2448,15 @@ bool SemaOpenACC::CheckVarIsPointerType(OpenACCClauseKind ClauseKind, ExprResult SemaOpenACC::ActOnVar(OpenACCClauseKind CK, Expr *VarExpr) { Expr *CurVarExpr = VarExpr->IgnoreParenImpCasts(); + // 'use_device' doesn't allow array subscript or array sections. + // OpenACC3.3 2.8: + // A 'var' in a 'use_device' clause must be the name of a variable or array. + if (CK == OpenACCClauseKind::UseDevice && + isa<ArraySectionExpr, ArraySubscriptExpr>(CurVarExpr)) { + Diag(VarExpr->getExprLoc(), diag::err_acc_not_a_var_ref_use_device); + return ExprError(); + } + // Sub-arrays/subscript-exprs are fine as long as the base is a // VarExpr/MemberExpr. So strip all of those off. while (isa<ArraySectionExpr, ArraySubscriptExpr>(CurVarExpr)) { @@ -2451,16 +2477,20 @@ ExprResult SemaOpenACC::ActOnVar(OpenACCClauseKind CK, Expr *VarExpr) { // If CK is a Reduction, this special cases for OpenACC3.3 2.5.15: "A var in a // reduction clause must be a scalar variable name, an aggregate variable // name, an array element, or a subarray. - // A MemberExpr that references a Field is valid. - if (CK != OpenACCClauseKind::Reduction) { + // If CK is a 'use_device', this also isn't valid, as it isn' the name of a + // variable or array. + // A MemberExpr that references a Field is valid for other clauses. + if (CK != OpenACCClauseKind::Reduction && + CK != OpenACCClauseKind::UseDevice) { if (const auto *ME = dyn_cast<MemberExpr>(CurVarExpr)) { if (isa<FieldDecl>(ME->getMemberDecl()->getCanonicalDecl())) return VarExpr; } } - // Referring to 'this' is always OK. - if (isa<CXXThisExpr>(CurVarExpr)) + // Referring to 'this' is ok for the most part, but for 'use_device' doesn't + // fall into 'variable or array name' + if (CK != OpenACCClauseKind::UseDevice && isa<CXXThisExpr>(CurVarExpr)) return VarExpr; // Nothing really we can do here, as these are dependent. So just return they @@ -2475,8 +2505,11 @@ ExprResult SemaOpenACC::ActOnVar(OpenACCClauseKind CK, Expr *VarExpr) { if (isa<RecoveryExpr>(CurVarExpr)) return ExprError(); - Diag(VarExpr->getExprLoc(), diag::err_acc_not_a_var_ref) - << (CK != OpenACCClauseKind::Reduction); + if (CK == OpenACCClauseKind::UseDevice) + Diag(VarExpr->getExprLoc(), diag::err_acc_not_a_var_ref_use_device); + else + Diag(VarExpr->getExprLoc(), diag::err_acc_not_a_var_ref) + << (CK != OpenACCClauseKind::Reduction); return ExprError(); } diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h index c33648ca0e34b8..c9bb079a9bcb34 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -11788,6 +11788,17 @@ void OpenACCClauseTransform<Derived>::VisitDeleteClause( ParsedClause.getEndLoc()); } +template <typename Derived> +void OpenACCClauseTransform<Derived>::VisitUseDeviceClause( + const OpenACCUseDeviceClause &C) { + ParsedClause.setVarListDetails(VisitVarList(C.getVarList()), + /*IsReadOnly=*/false, /*IsZero=*/false); + NewClause = OpenACCUseDeviceClause::Create( + Self.getSema().getASTContext(), ParsedClause.getBeginLoc(), + ParsedClause.getLParenLoc(), ParsedClause.getVarList(), + ParsedClause.getEndLoc()); +} + template <typename Derived> void OpenACCClauseTransform<Derived>::VisitDevicePtrClause( const OpenACCDevicePtrClause &C) { diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp index 741bae684cffe3..21f6b2ecc58c4f 100644 --- a/clang/lib/Serialization/ASTReader.cpp +++ b/clang/lib/Serialization/ASTReader.cpp @@ -12441,6 +12441,12 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { return OpenACCDeleteClause::Create(getContext(), BeginLoc, LParenLoc, VarList, EndLoc); } + case OpenACCClauseKind::UseDevice: { + SourceLocation LParenLoc = readSourceLocation(); + llvm::SmallVector<Expr *> VarList = readOpenACCVarList(); + return OpenACCUseDeviceClause::Create(getContext(), BeginLoc, LParenLoc, + VarList, EndLoc); + } case OpenACCClauseKind::DevicePtr: { SourceLocation LParenLoc = readSourceLocation(); llvm::SmallVector<Expr *> VarList = readOpenACCVarList(); @@ -12583,7 +12589,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { } case OpenACCClauseKind::NoHost: - case OpenACCClauseKind::UseDevice: case OpenACCClauseKind::Device: case OpenACCClauseKind::DeviceResident: case OpenACCClauseKind::Host: diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp index 9517bad4070dfb..6db2262a7952ec 100644 --- a/clang/lib/Serialization/ASTWriter.cpp +++ b/clang/lib/Serialization/ASTWriter.cpp @@ -8368,6 +8368,12 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { writeOpenACCVarList(DC); return; } + case OpenACCClauseKind::UseDevice: { + const auto *UDC = cast<OpenACCUseDeviceClause>(C); + writeSourceLocation(UDC->getLParenLoc()); + writeOpenACCVarList(UDC); + return; + } case OpenACCClauseKind::DevicePtr: { const auto *DPC = cast<OpenACCDevicePtrClause>(C); writeSourceLocation(DPC->getLParenLoc()); @@ -8511,7 +8517,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { } case OpenACCClauseKind::NoHost: - case OpenACCClauseKind::UseDevice: case OpenACCClauseKind::Device: case OpenACCClauseKind::DeviceResident: case OpenACCClauseKind::Host: diff --git a/clang/test/AST/ast-print-openacc-data-construct.cpp b/clang/test/AST/ast-print-openacc-data-construct.cpp index f03f96239ab4c6..7ad0a43e322a21 100644 --- a/clang/test/AST/ast-print-openacc-data-construct.cpp +++ b/clang/test/AST/ast-print-openacc-data-construct.cpp @@ -20,8 +20,7 @@ void foo() { // CHECK: #pragma acc exit data copyout(Var) #pragma acc exit data copyout(Var) ; -// CHECK: #pragma acc host_data -// CHECK-NOT: use_device(Var) +// CHECK: #pragma acc host_data use_device(Var) #pragma acc host_data use_device(Var) ; @@ -38,7 +37,7 @@ void foo() { // CHECK: #pragma acc exit data copyout(Var) if(i == array[1]) #pragma acc exit data copyout(Var) if(i == array[1]) ; -// CHECK: #pragma acc host_data if(i == array[1]) +// CHECK: #pragma acc host_data use_device(Var) if(i == array[1]) #pragma acc host_data use_device(Var) if(i == array[1]) ; @@ -114,7 +113,7 @@ void foo() { // CHECK: #pragma acc exit data copyout(i) finalize #pragma acc exit data copyout(i) finalize -// CHECK: #pragma acc host_data if_present +// CHECK: #pragma acc host_data use_device(i) if_present #pragma acc host_data use_device(i) if_present ; // CHECK: #pragma acc exit data copyout(i) detach(iPtr, arrayPtr[0]) @@ -126,4 +125,8 @@ void foo() { // CHECK: #pragma acc exit data copyout(i) delete(i, array[1], array, array[1:2]) #pragma acc exit data copyout(i) delete(i, array[1], array, array[1:2]) + +// CHECK: #pragma acc host_data use_device(i) +#pragma acc host_data use_device(i) + ; } diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c index e583fb3897998d..487dc79f538085 100644 --- a/clang/test/ParserOpenACC/parse-clauses.c +++ b/clang/test/ParserOpenACC/parse-clauses.c @@ -443,13 +443,16 @@ void VarListClauses() { #pragma acc serial present_or_copy(HasMem.MemArr[3:]) for(int i = 0; i < 5;++i) {} - // expected-error@+2{{expected ','}} - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented, clause ignored}} -#pragma acc serial use_device(s.array[s.value] s.array[s.value :5] ), self + // expected-error@+2 2{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} + // expected-error@+1{{expected ','}} +#pragma acc host_data use_device(s.array[s.value] s.array[s.value :5] ), if_present for(int i = 0; i < 5;++i) {} - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented, clause ignored}} -#pragma acc serial use_device(s.array[s.value : 5]), self + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device(s.array[s.value : 5]), if_present + for(int i = 0; i < 5;++i) {} + +#pragma acc host_data use_device(HasMem), if_present for(int i = 0; i < 5;++i) {} // expected-error@+1{{expected ','}} @@ -517,15 +520,6 @@ void VarListClauses() { #pragma acc exit data delete(s.array[s.value : 5], s.value),async for(int i = 0; i < 5;++i) {} - // expected-error@+2{{expected ','}} - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented, clause ignored}} -#pragma acc exit data use_device(s.array[s.value] s.array[s.value :5] ),async - for(int i = 0; i < 5;++i) {} - - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented, clause ignored}} -#pragma acc exit data use_device(s.array[s.value : 5], s.value), async - for(int i = 0; i < 5;++i) {} - // expected-error@+2{{expected ','}} // expected-warning@+1{{OpenACC clause 'device_resident' not yet implemented, clause ignored}} #pragma acc serial device_resident(s.array[s.value] s.array[s.value :5] ), self 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 a9f6f1e6b9e3a3..a71d25ce61bed4 100644 --- a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c +++ b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c @@ -64,7 +64,7 @@ void uses() { // expected-warning@+1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}} #pragma acc parallel loop auto present_or_copy(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'parallel loop' directive}} #pragma acc parallel loop auto use_device(Var) for(unsigned i = 0; i < 5; ++i); #pragma acc parallel loop auto attach(VarPtr) @@ -181,7 +181,7 @@ void uses() { // expected-warning@+1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}} #pragma acc parallel loop present_or_copy(Var) auto for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'parallel loop' directive}} #pragma acc parallel loop use_device(Var) auto for(unsigned i = 0; i < 5; ++i); #pragma acc parallel loop attach(VarPtr) auto @@ -299,7 +299,7 @@ void uses() { // expected-warning@+1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}} #pragma acc parallel loop independent present_or_copy(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'parallel loop' directive}} #pragma acc parallel loop independent use_device(Var) for(unsigned i = 0; i < 5; ++i); #pragma acc parallel loop independent attach(VarPtr) @@ -416,7 +416,7 @@ void uses() { // expected-warning@+1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}} #pragma acc parallel loop present_or_copy(Var) independent for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'parallel loop' directive}} #pragma acc parallel loop use_device(Var) independent for(unsigned i = 0; i < 5; ++i); #pragma acc parallel loop attach(VarPtr) independent @@ -542,7 +542,7 @@ void uses() { // expected-warning@+1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}} #pragma acc parallel loop seq present_or_copy(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'parallel loop' directive}} #pragma acc parallel loop seq use_device(Var) for(unsigned i = 0; i < 5; ++i); #pragma acc parallel loop seq attach(VarPtr) @@ -665,7 +665,7 @@ void uses() { // expected-warning@+1{{OpenACC clause name 'present_or_copy' is a deprecated clause name and is now an alias for 'copy'}} #pragma acc parallel loop present_or_copy(Var) seq for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'parallel loop' directive}} #pragma acc parallel loop use_device(Var) seq for(unsigned i = 0; i < 5; ++i); #pragma acc parallel loop attach(VarPtr) seq diff --git a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c index 4526a11eeb9079..40339941f51a9c 100644 --- a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c +++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c @@ -87,8 +87,7 @@ void uses() { // expected-note@+1{{previous clause is here}} #pragma acc serial loop device_type(*) present_or_copy(Var) for(int i = 0; i < 5; ++i); - // expected-error@+2{{OpenACC clause 'use_device' may not follow a 'device_type' clause in a 'kernels loop' construct}} - // expected-note@+1{{previous clause is here}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'kernels loop' directive}} #pragma acc kernels loop device_type(*) use_device(Var) for(int i = 0; i < 5; ++i); // expected-error@+2{{OpenACC clause 'attach' 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 6f46e615f43c9d..cccb50ee55582d 100644 --- a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c @@ -91,8 +91,7 @@ void uses() { // expected-note@+1{{previous clause is here}} #pragma acc kernels device_type(*) present_or_copy(Var) while(1); - // expected-error@+2{{OpenACC clause 'use_device' may not follow a 'device_type' clause in a 'kernels' construct}} - // expected-note@+1{{previous clause is here}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'kernels' directive}} #pragma acc kernels device_type(*) use_device(Var) while(1); // expected-error@+2{{OpenACC clause 'attach' may not follow a 'device_type' clause in a 'kernels' construct}} diff --git a/clang/test/SemaOpenACC/data-construct-ast.cpp b/clang/test/SemaOpenACC/data-construct-ast.cpp index f299fd04581a72..5abd142b237a21 100644 --- a/clang/test/SemaOpenACC/data-construct-ast.cpp +++ b/clang/test/SemaOpenACC/data-construct-ast.cpp @@ -36,6 +36,8 @@ void NormalFunc() { #pragma acc host_data use_device(Var) while (Var); // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Var' 'int' // CHECK-NEXT: WhileStmt // CHECK: NullStmt } @@ -68,6 +70,8 @@ void TemplFunc() { #pragma acc host_data use_device(Var) while (Var); // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Var' 'T' // CHECK-NEXT: WhileStmt // CHECK: NullStmt @@ -94,6 +98,8 @@ void TemplFunc() { // CHECK-NEXT: DeclRefExpr{{.*}}'Var' 'int' // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Var' 'int' // CHECK-NEXT: WhileStmt // CHECK: NullStmt } diff --git a/clang/test/SemaOpenACC/data-construct-async-clause.c b/clang/test/SemaOpenACC/data-construct-async-clause.c index 7173b2f0be7dd9..3c9fbae0d9875d 100644 --- a/clang/test/SemaOpenACC/data-construct-async-clause.c +++ b/clang/test/SemaOpenACC/data-construct-async-clause.c @@ -9,7 +9,6 @@ void Test() { ; #pragma acc enter data copyin(I) async(I) #pragma acc exit data copyout(I) async(I) - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC 'async' clause is not valid on 'host_data' directive}} #pragma acc host_data use_device(I) async(I) ; @@ -21,7 +20,6 @@ void Test() { #pragma acc enter data copyin(NC) async(NC) // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('struct NotConvertible' invalid)}} #pragma acc exit data copyout(NC) async(NC) - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC clause 'async' requires expression of integer type ('struct NotConvertible' invalid)}} #pragma acc host_data use_device(NC) async(NC) ; diff --git a/clang/test/SemaOpenACC/data-construct-attach-clause.c b/clang/test/SemaOpenACC/data-construct-attach-clause.c index 49a708e49d24b2..0bc02563ff6953 100644 --- a/clang/test/SemaOpenACC/data-construct-attach-clause.c +++ b/clang/test/SemaOpenACC/data-construct-attach-clause.c @@ -58,7 +58,6 @@ void uses() { // expected-error@+1{{OpenACC 'attach' clause is not valid on 'exit data' directive}} #pragma acc exit data copyout(LocalInt) attach(PtrArray[0]) - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC 'attach' clause is not valid on 'host_data' directive}} #pragma acc host_data use_device(LocalInt) attach(PtrArray[0]) ; diff --git a/clang/test/SemaOpenACC/data-construct-detach-clause.c b/clang/test/SemaOpenACC/data-construct-detach-clause.c index e75c95d99ec078..edcc80ac362362 100644 --- a/clang/test/SemaOpenACC/data-construct-detach-clause.c +++ b/clang/test/SemaOpenACC/data-construct-detach-clause.c @@ -61,7 +61,6 @@ void uses() { ; // expected-error@+1{{OpenACC 'detach' clause is not valid on 'enter data' directive}} #pragma acc enter data copyin(LocalInt) detach(PtrArray[0]) - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC 'detach' clause is not valid on 'host_data' directive}} #pragma acc host_data use_device(LocalInt) detach(PtrArray[0]) ; diff --git a/clang/test/SemaOpenACC/data-construct-deviceptr-clause.c b/clang/test/SemaOpenACC/data-construct-deviceptr-clause.c index d7869e965c5b6a..70ccd3b7aec953 100644 --- a/clang/test/SemaOpenACC/data-construct-deviceptr-clause.c +++ b/clang/test/SemaOpenACC/data-construct-deviceptr-clause.c @@ -61,7 +61,6 @@ void uses() { #pragma acc enter data copyin(LocalInt) deviceptr(LocalInt) // expected-error@+1{{OpenACC 'deviceptr' clause is not valid on 'exit data' directive}} #pragma acc exit data copyout(LocalInt) deviceptr(LocalInt) - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC 'deviceptr' clause is not valid on 'host_data' directive}} #pragma acc host_data use_device(LocalInt) deviceptr(LocalInt) ; diff --git a/clang/test/SemaOpenACC/data-construct-finalize-clause.c b/clang/test/SemaOpenACC/data-construct-finalize-clause.c index b2b4ada0e42ed9..252b26708cd811 100644 --- a/clang/test/SemaOpenACC/data-construct-finalize-clause.c +++ b/clang/test/SemaOpenACC/data-construct-finalize-clause.c @@ -13,7 +13,6 @@ void Test() { // finalize is valid only on exit data, otherwise has no other rules. #pragma acc exit data copyout(I) finalize ; - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC 'finalize' clause is not valid on 'host_data' directive}} #pragma acc host_data use_device(I) finalize ; diff --git a/clang/test/SemaOpenACC/data-construct-if-ast.cpp b/clang/test/SemaOpenACC/data-construct-if-ast.cpp index 3b810a724c51ee..9ceee4e1c07497 100644 --- a/clang/test/SemaOpenACC/data-construct-if-ast.cpp +++ b/clang/test/SemaOpenACC/data-construct-if-ast.cpp @@ -72,6 +72,8 @@ void TemplFunc() { #pragma acc host_data use_device(Global) if(T::BC) ; // CHECK-NEXT: OpenACCHostDataConstruct{{.*}}host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Global' 'int' // CHECK-NEXT: if clause // CHECK-NEXT: DependentScopeDeclRefExpr{{.*}} '<dependent type>' lvalue // CHECK-NEXT: NestedNameSpecifier TypeSpec 'T' @@ -116,6 +118,8 @@ void TemplFunc() { // CHECK-NEXT: NullStmt // CHECK-NEXT: OpenACCHostDataConstruct{{.*}}host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Global' 'int' // CHECK-NEXT: if clause // CHECK-NEXT: ImplicitCastExpr{{.*}} 'bool' <UserDefinedConversion> // CHECK-NEXT: CXXMemberCallExpr{{.*}} 'bool' diff --git a/clang/test/SemaOpenACC/data-construct-if-clause.c b/clang/test/SemaOpenACC/data-construct-if-clause.c index 0a7989e80fc373..f22452d2c34a41 100644 --- a/clang/test/SemaOpenACC/data-construct-if-clause.c +++ b/clang/test/SemaOpenACC/data-construct-if-clause.c @@ -20,10 +20,8 @@ void Foo() { // expected-note@+1{{previous clause is here}} #pragma acc exit data copyout(Var) if(1) if (2) - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} #pragma acc host_data use_device(Var) if(1) ; - // expected-warning@+3{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+2{{OpenACC 'if' clause cannot appear more than once on a 'host_data' directive}} // expected-note@+1{{previous clause is here}} #pragma acc host_data use_device(Var) if(1) if (2) diff --git a/clang/test/SemaOpenACC/data-construct-if_present-ast.cpp b/clang/test/SemaOpenACC/data-construct-if_present-ast.cpp index d2ef6822819655..8dab6bcce58d9c 100644 --- a/clang/test/SemaOpenACC/data-construct-if_present-ast.cpp +++ b/clang/test/SemaOpenACC/data-construct-if_present-ast.cpp @@ -17,6 +17,8 @@ void Uses() { #pragma acc host_data use_device(I) if_present ; // CHECK-NEXT: OpenACCHostDataConstruct{{.*}}host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'I' 'int' // CHECK-NEXT: if_present clause // CHECK-NEXT: NullStmt } @@ -35,6 +37,8 @@ void TemplUses() { #pragma acc host_data use_device(I) if_present ; // CHECK-NEXT: OpenACCHostDataConstruct{{.*}}host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'I' 'T' // CHECK-NEXT: if_present clause // CHECK-NEXT: NullStmt @@ -48,6 +52,8 @@ void TemplUses() { // CHECK-NEXT: VarDecl // CHECK-NEXT: OpenACCHostDataConstruct{{.*}}host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'I' 'int' // CHECK-NEXT: if_present clause // CHECK-NEXT: NullStmt } diff --git a/clang/test/SemaOpenACC/data-construct-if_present-clause.c b/clang/test/SemaOpenACC/data-construct-if_present-clause.c index ce92ec024b4f99..b1290cdccca5f1 100644 --- a/clang/test/SemaOpenACC/data-construct-if_present-clause.c +++ b/clang/test/SemaOpenACC/data-construct-if_present-clause.c @@ -13,7 +13,6 @@ void Test() { // expected-error@+1{{OpenACC 'if_present' clause is not valid on 'exit data' directive}} #pragma acc exit data copyout(I) if_present ; - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} #pragma acc host_data use_device(I) if_present ; } diff --git a/clang/test/SemaOpenACC/data-construct-present-clause.c b/clang/test/SemaOpenACC/data-construct-present-clause.c index 3128f532faf3f0..b889230d177cdc 100644 --- a/clang/test/SemaOpenACC/data-construct-present-clause.c +++ b/clang/test/SemaOpenACC/data-construct-present-clause.c @@ -52,7 +52,6 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo #pragma acc enter data copyin(LocalInt) present(LocalInt) // expected-error@+1{{OpenACC 'present' clause is not valid on 'exit data' directive}} #pragma acc exit data copyout(LocalInt) present(LocalInt) - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC 'present' clause is not valid on 'host_data' directive}} #pragma acc host_data use_device(LocalInt) present(LocalInt) ; diff --git a/clang/test/SemaOpenACC/data-construct-use_device-ast.cpp b/clang/test/SemaOpenACC/data-construct-use_device-ast.cpp new file mode 100644 index 00000000000000..b8cc30e14671fa --- /dev/null +++ b/clang/test/SemaOpenACC/data-construct-use_device-ast.cpp @@ -0,0 +1,58 @@ +// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s + +// Test this with PCH. +// RUN: %clang_cc1 %s -fopenacc -emit-pch -o %t %s +// RUN: %clang_cc1 %s -fopenacc -include-pch %t -ast-dump-all | FileCheck %s + +#ifndef PCH_HELPER +#define PCH_HELPER + +int Global; +short GlobalArray[5]; +void NormalUses(float *PointerParam) { + // CHECK: FunctionDecl{{.*}}NormalUses + // CHECK: ParmVarDecl + // CHECK-NEXT: CompoundStmt + +#pragma acc host_data use_device(GlobalArray, PointerParam) + ; + // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]' + // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *' + // CHECK-NEXT: NullStmt +} + +template<typename T> +void TemplUses(T t) { + // CHECK-NEXT: FunctionTemplateDecl + // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 0 T + // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void (T)' + // CHECK-NEXT: ParmVarDecl{{.*}} referenced t 'T' + // CHECK-NEXT: CompoundStmt + +#pragma acc host_data use_device(t) + ; + // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T' + // CHECK-NEXT: NullStmt + + // Check the instantiated versions of the above. + // CHECK-NEXT: FunctionDecl{{.*}} used TemplUses 'void (int)' implicit_instantiation + // CHECK-NEXT: TemplateArgument type 'int' + // CHECK-NEXT: BuiltinType{{.*}} 'int' + // CHECK-NEXT: ParmVarDecl{{.*}} used t 'int' + // CHECK-NEXT: CompoundStmt + + // CHECK-NEXT: OpenACCHostDataConstruct{{.*}} host_data + // CHECK-NEXT: use_device clause + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int' + // CHECK-NEXT: NullStmt +} + +void Inst() { + int i; + TemplUses(i); +} +#endif diff --git a/clang/test/SemaOpenACC/data-construct-use_device-clause.c b/clang/test/SemaOpenACC/data-construct-use_device-clause.c new file mode 100644 index 00000000000000..d0f74585759cff --- /dev/null +++ b/clang/test/SemaOpenACC/data-construct-use_device-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 host_data use_device(LocalInt) + + // Valid cases: +#pragma acc host_data use_device(LocalInt, LocalPointer, LocalArray) + ; + ; + // expected-error@+2{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device(LocalComposite.ScalarMember, LocalComposite.ScalarMember) + ; + + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device(LocalArray[2:1]) + + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device(1 + IntParam) + ; + + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device(+IntParam) + ; + + // expected-error@+2{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device(PointerParam[2:]) + ; + + // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device(ArrayParam[2:5]) + ; + + // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device((float*)ArrayParam[2:5]) + ; + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} +#pragma acc host_data use_device((float)ArrayParam[2]) + ; + + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'data' directive}} +#pragma acc data use_device(LocalInt) + ; + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'enter data' directive}} +#pragma acc enter data use_device(LocalInt) + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'exit data' directive}} +#pragma acc exit data use_device(LocalInt) +} diff --git a/clang/test/SemaOpenACC/data-construct-wait-clause.c b/clang/test/SemaOpenACC/data-construct-wait-clause.c index dffcba34e7b333..cef2dbdca29ed8 100644 --- a/clang/test/SemaOpenACC/data-construct-wait-clause.c +++ b/clang/test/SemaOpenACC/data-construct-wait-clause.c @@ -14,9 +14,8 @@ void uses() { #pragma acc exit data copyout(arr[0]) wait(getS(), getI()) - // expected-warning@+2{{OpenACC clause 'use_device' not yet implemented}} // expected-error@+1{{OpenACC 'wait' clause is not valid on 'host_data' directive}} -#pragma acc host_data use_device(arr[0]) wait(getS(), getI()) +#pragma acc host_data use_device(arr) wait(getS(), getI()) ; #pragma acc data copyin(arr[0]) wait(devnum:getS(): getI()) diff --git a/clang/test/SemaOpenACC/data-construct.cpp b/clang/test/SemaOpenACC/data-construct.cpp index 507cc5ac5cfaf5..309d0300d5a9e1 100644 --- a/clang/test/SemaOpenACC/data-construct.cpp +++ b/clang/test/SemaOpenACC/data-construct.cpp @@ -86,7 +86,6 @@ void AtLeastOneOf() { #pragma acc exit data // Host Data - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} #pragma acc host_data use_device(Var) ; // OpenACC TODO: The following 'host_data' directives should diagnose, since @@ -102,9 +101,6 @@ void AtLeastOneOf() { void DataRules() { int Var; - // OpenACC TODO: Only 'async' and 'wait' are permitted after a device_type, so - // the rest of these should diagnose. - // expected-error@+2{{OpenACC clause 'copy' may not follow a 'device_type' clause in a 'data' construct}} // expected-note@+1{{previous clause is here}} #pragma acc data device_type(*) copy(Var) @@ -155,16 +151,13 @@ struct HasMembers { int Member; void HostDataError() { - // TODO OpenACC: The following 3 should error, as use_device's var only allows - // a variable or array, not an array index, or sub expression. - - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} #pragma acc host_data use_device(this) ; - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} #pragma acc host_data use_device(this->Member) ; - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} #pragma acc host_data use_device(Member) ; } @@ -177,27 +170,22 @@ void HostDataRules() { #pragma acc host_data if(Var) if (Var2) ; - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} #pragma acc host_data use_device(Var) ; int Array[5]; - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} #pragma acc host_data use_device(Array) ; - // TODO OpenACC: The following 3 should error, as use_device's var only allows - // a variable or array, not an array index, or sub expression. - - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} #pragma acc host_data use_device(Array[1:1]) ; - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} #pragma acc host_data use_device(Array[1]) ; HasMembers HM; - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC variable in 'use_device' clause is not a valid variable name or array name}} #pragma acc host_data use_device(HM.Member) ; 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 bbd04e7afa6f25..d196633c8b6d92 100644 --- a/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c +++ b/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c @@ -68,7 +68,7 @@ void uses() { // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop auto present_or_copy(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'loop' directive}} #pragma acc loop auto use_device(Var) for(unsigned i = 0; i < 5; ++i); // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} @@ -202,7 +202,7 @@ void uses() { // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop present_or_copy(Var) auto for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'loop' directive}} #pragma acc loop use_device(Var) auto for(unsigned i = 0; i < 5; ++i); // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} @@ -337,7 +337,7 @@ void uses() { // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop independent present_or_copy(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'loop' directive}} #pragma acc loop independent use_device(Var) for(unsigned i = 0; i < 5; ++i); // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} @@ -471,7 +471,7 @@ void uses() { // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop present_or_copy(Var) independent for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'loop' directive}} #pragma acc loop use_device(Var) independent for(unsigned i = 0; i < 5; ++i); // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} @@ -614,7 +614,7 @@ void uses() { // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop seq present_or_copy(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'loop' directive}} #pragma acc loop seq use_device(Var) for(unsigned i = 0; i < 5; ++i); // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} @@ -754,7 +754,7 @@ void uses() { // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop present_or_copy(Var) seq for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'use_device' not yet implemented}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'loop' directive}} #pragma acc loop use_device(Var) seq for(unsigned i = 0; i < 5; ++i); // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} diff --git a/clang/test/SemaOpenACC/loop-construct-device_type-clause.c b/clang/test/SemaOpenACC/loop-construct-device_type-clause.c index 8fc6690273c7de..3e4d0da60b6b27 100644 --- a/clang/test/SemaOpenACC/loop-construct-device_type-clause.c +++ b/clang/test/SemaOpenACC/loop-construct-device_type-clause.c @@ -80,8 +80,7 @@ void uses() { // expected-error@+1{{OpenACC 'present_or_copy' clause is not valid on 'loop' directive}} #pragma acc loop device_type(*) present_or_copy(Var) for(int i = 0; i < 5; ++i); - // expected-error@+2{{OpenACC clause 'use_device' may not follow a 'device_type' clause in a 'loop' construct}} - // expected-note@+1{{previous clause is here}} + // expected-error@+1{{OpenACC 'use_device' clause is not valid on 'loop' directive}} #pragma acc loop device_type(*) use_device(Var) for(int i = 0; i < 5; ++i); // expected-error@+1{{OpenACC 'attach' clause is not valid on 'loop' directive}} diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp index d1a28624618990..701582138e053d 100644 --- a/clang/tools/libclang/CIndex.cpp +++ b/clang/tools/libclang/CIndex.cpp @@ -2894,6 +2894,11 @@ void OpenACCClauseEnqueue::VisitDeleteClause(const OpenACCDeleteClause &C) { VisitVarList(C); } +void OpenACCClauseEnqueue::VisitUseDeviceClause( + const OpenACCUseDeviceClause &C) { + VisitVarList(C); +} + void OpenACCClauseEnqueue::VisitDevicePtrClause( const OpenACCDevicePtrClause &C) { VisitVarList(C); _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
