Author: erichkeane Date: 2024-12-13T13:51:41-08:00 New Revision: 3351b3bf8dcb9aebfa6f491fcbe5a00acbcc3291
URL: https://github.com/llvm/llvm-project/commit/3351b3bf8dcb9aebfa6f491fcbe5a00acbcc3291 DIFF: https://github.com/llvm/llvm-project/commit/3351b3bf8dcb9aebfa6f491fcbe5a00acbcc3291.diff LOG: [OpenACC] implement 'detach' clause sema This is another new clause specific to 'exit data' that takes a pointer argument. This patch implements this the same way we do a few other clauses (like attach) that have the same restrictions. Added: clang/test/SemaOpenACC/data-construct-detach-ast.cpp clang/test/SemaOpenACC/data-construct-detach-clause.c 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-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.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 57cd9a1e8564d2..7778f8199b3af6 100644 --- a/clang/include/clang/AST/OpenACCClause.h +++ b/clang/include/clang/AST/OpenACCClause.h @@ -744,6 +744,28 @@ class OpenACCAttachClause final ArrayRef<Expr *> VarList, SourceLocation EndLoc); }; +class OpenACCDetachClause final + : public OpenACCClauseWithVarList, + public llvm::TrailingObjects<OpenACCDetachClause, Expr *> { + + OpenACCDetachClause(SourceLocation BeginLoc, SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, SourceLocation EndLoc) + : OpenACCClauseWithVarList(OpenACCClauseKind::Detach, 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::Detach; + } + static OpenACCDetachClause * + Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, SourceLocation EndLoc); +}; + class OpenACCNoCreateClause final : public OpenACCClauseWithVarList, public llvm::TrailingObjects<OpenACCNoCreateClause, Expr *> { diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def index cc788e4de45279..87983c3849480f 100644 --- a/clang/include/clang/Basic/OpenACCClauses.def +++ b/clang/include/clang/Basic/OpenACCClauses.def @@ -38,6 +38,7 @@ VISIT_CLAUSE(Create) CLAUSE_ALIAS(PCreate, Create, true) CLAUSE_ALIAS(PresentOrCreate, Create, true) VISIT_CLAUSE(Default) +VISIT_CLAUSE(Detach) VISIT_CLAUSE(DevicePtr) VISIT_CLAUSE(DeviceType) CLAUSE_ALIAS(DType, DeviceType, false) diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h index 170a6f4885c965..ea3f34e3f4a959 100644 --- a/clang/include/clang/Sema/SemaOpenACC.h +++ b/clang/include/clang/Sema/SemaOpenACC.h @@ -399,6 +399,7 @@ class SemaOpenACC : public SemaBase { ClauseKind == OpenACCClauseKind::PCreate || ClauseKind == OpenACCClauseKind::PresentOrCreate || ClauseKind == OpenACCClauseKind::Attach || + ClauseKind == OpenACCClauseKind::Detach || ClauseKind == OpenACCClauseKind::DevicePtr || ClauseKind == OpenACCClauseKind::Reduction || ClauseKind == OpenACCClauseKind::FirstPrivate) && @@ -535,6 +536,7 @@ class SemaOpenACC : public SemaBase { ClauseKind == OpenACCClauseKind::PCreate || ClauseKind == OpenACCClauseKind::PresentOrCreate || ClauseKind == OpenACCClauseKind::Attach || + ClauseKind == OpenACCClauseKind::Detach || ClauseKind == OpenACCClauseKind::DevicePtr || ClauseKind == OpenACCClauseKind::FirstPrivate) && "Parsed clause kind does not have a var-list"); @@ -571,6 +573,7 @@ class SemaOpenACC : public SemaBase { ClauseKind == OpenACCClauseKind::PCreate || ClauseKind == OpenACCClauseKind::PresentOrCreate || ClauseKind == OpenACCClauseKind::Attach || + ClauseKind == OpenACCClauseKind::Detach || ClauseKind == OpenACCClauseKind::DevicePtr || ClauseKind == OpenACCClauseKind::FirstPrivate) && "Parsed clause kind does not have a var-list"); diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp index 443cfa84474bed..d2d8f34e9014de 100644 --- a/clang/lib/AST/OpenACCClause.cpp +++ b/clang/lib/AST/OpenACCClause.cpp @@ -33,7 +33,8 @@ bool OpenACCClauseWithVarList::classof(const OpenACCClause *C) { OpenACCFirstPrivateClause::classof(C) || OpenACCDevicePtrClause::classof(C) || OpenACCDevicePtrClause::classof(C) || - OpenACCAttachClause::classof(C) || OpenACCNoCreateClause::classof(C) || + OpenACCDetachClause::classof(C) || OpenACCAttachClause::classof(C) || + OpenACCNoCreateClause::classof(C) || OpenACCPresentClause::classof(C) || OpenACCCopyClause::classof(C) || OpenACCCopyInClause::classof(C) || OpenACCCopyOutClause::classof(C) || OpenACCReductionClause::classof(C) || OpenACCCreateClause::classof(C); @@ -277,6 +278,16 @@ OpenACCAttachClause *OpenACCAttachClause::Create(const ASTContext &C, return new (Mem) OpenACCAttachClause(BeginLoc, LParenLoc, VarList, EndLoc); } +OpenACCDetachClause *OpenACCDetachClause::Create(const ASTContext &C, + SourceLocation BeginLoc, + SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, + SourceLocation EndLoc) { + void *Mem = + C.Allocate(OpenACCDetachClause::totalSizeToAlloc<Expr *>(VarList.size())); + return new (Mem) OpenACCDetachClause(BeginLoc, LParenLoc, VarList, EndLoc); +} + OpenACCDevicePtrClause *OpenACCDevicePtrClause::Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc, @@ -546,6 +557,13 @@ void OpenACCClausePrinter::VisitAttachClause(const OpenACCAttachClause &C) { OS << ")"; } +void OpenACCClausePrinter::VisitDetachClause(const OpenACCDetachClause &C) { + OS << "detach("; + 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 bd4956c15eea1c..4e9865f722f78e 100644 --- a/clang/lib/AST/StmtProfile.cpp +++ b/clang/lib/AST/StmtProfile.cpp @@ -2605,6 +2605,12 @@ void OpenACCClauseProfiler::VisitAttachClause( Profiler.VisitStmt(E); } +void OpenACCClauseProfiler::VisitDetachClause( + const OpenACCDetachClause &Clause) { + for (auto *E : Clause.getVarList()) + Profiler.VisitStmt(E); +} + void OpenACCClauseProfiler::VisitDevicePtrClause( const OpenACCDevicePtrClause &Clause) { for (auto *E : Clause.getVarList()) diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp index 0a5c0d561203af..b02b682fb0c58f 100644 --- a/clang/lib/AST/TextNodeDumper.cpp +++ b/clang/lib/AST/TextNodeDumper.cpp @@ -411,6 +411,7 @@ void TextNodeDumper::Visit(const OpenACCClause *C) { case OpenACCClauseKind::If: case OpenACCClauseKind::IfPresent: case OpenACCClauseKind::Independent: + case OpenACCClauseKind::Detach: case OpenACCClauseKind::DevicePtr: case OpenACCClauseKind::Finalize: case OpenACCClauseKind::FirstPrivate: diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp index 8c81936b35296c..5da7069edaa740 100644 --- a/clang/lib/Parse/ParseOpenACC.cpp +++ b/clang/lib/Parse/ParseOpenACC.cpp @@ -999,7 +999,6 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams( assert(DirKind == OpenACCDirectiveKind::Update); [[fallthrough]]; case OpenACCClauseKind::Delete: - case OpenACCClauseKind::Detach: case OpenACCClauseKind::Device: case OpenACCClauseKind::DeviceResident: case OpenACCClauseKind::Host: @@ -1008,6 +1007,7 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams( ParseOpenACCVarList(ClauseKind); break; case OpenACCClauseKind::Attach: + case OpenACCClauseKind::Detach: case OpenACCClauseKind::DevicePtr: ParsedClause.setVarListDetails(ParseOpenACCVarList(ClauseKind), /*IsReadOnly=*/false, /*IsZero=*/false); diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp index 3917c9bf60b8c4..7156e37991284f 100644 --- a/clang/lib/Sema/SemaOpenACC.cpp +++ b/clang/lib/Sema/SemaOpenACC.cpp @@ -425,6 +425,14 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind, return false; } } + case OpenACCClauseKind::Detach: { + switch (DirectiveKind) { + case OpenACCDirectiveKind::ExitData: + return true; + default: + return false; + } + } } default: @@ -1043,6 +1051,21 @@ OpenACCClause *SemaOpenACCClauseVisitor::VisitAttachClause( Clause.getEndLoc()); } +OpenACCClause *SemaOpenACCClauseVisitor::VisitDetachClause( + SemaOpenACC::OpenACCParsedClause &Clause) { + // ActOnVar ensured that everything is a valid variable reference, but we + // still have to make sure it is a pointer type. + llvm::SmallVector<Expr *> VarList{Clause.getVarList()}; + llvm::erase_if(VarList, [&](Expr *E) { + return SemaRef.CheckVarIsPointerType(OpenACCClauseKind::Detach, E); + }); + Clause.setVarListDetails(VarList, + /*IsReadOnly=*/false, /*IsZero=*/false); + return OpenACCDetachClause::Create(Ctx, Clause.getBeginLoc(), + Clause.getLParenLoc(), Clause.getVarList(), + Clause.getEndLoc()); +} + OpenACCClause *SemaOpenACCClauseVisitor::VisitDevicePtrClause( 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 e5c3584b4d8820..0f92186d08933c 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -11755,6 +11755,28 @@ void OpenACCClauseTransform<Derived>::VisitAttachClause( ParsedClause.getEndLoc()); } +template <typename Derived> +void OpenACCClauseTransform<Derived>::VisitDetachClause( + const OpenACCDetachClause &C) { + llvm::SmallVector<Expr *> VarList = VisitVarList(C.getVarList()); + + // Ensure each var is a pointer type. + VarList.erase( + std::remove_if(VarList.begin(), VarList.end(), + [&](Expr *E) { + return Self.getSema().OpenACC().CheckVarIsPointerType( + OpenACCClauseKind::Detach, E); + }), + VarList.end()); + + ParsedClause.setVarListDetails(VarList, + /*IsReadOnly=*/false, /*IsZero=*/false); + NewClause = OpenACCDetachClause::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 4f7dc597d7bd5a..bed674233deecf 100644 --- a/clang/lib/Serialization/ASTReader.cpp +++ b/clang/lib/Serialization/ASTReader.cpp @@ -12442,6 +12442,12 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { return OpenACCAttachClause::Create(getContext(), BeginLoc, LParenLoc, VarList, EndLoc); } + case OpenACCClauseKind::Detach: { + SourceLocation LParenLoc = readSourceLocation(); + llvm::SmallVector<Expr *> VarList = readOpenACCVarList(); + return OpenACCDetachClause::Create(getContext(), BeginLoc, LParenLoc, + VarList, EndLoc); + } case OpenACCClauseKind::DevicePtr: { SourceLocation LParenLoc = readSourceLocation(); llvm::SmallVector<Expr *> VarList = readOpenACCVarList(); @@ -12586,7 +12592,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { case OpenACCClauseKind::NoHost: case OpenACCClauseKind::UseDevice: case OpenACCClauseKind::Delete: - case OpenACCClauseKind::Detach: case OpenACCClauseKind::Device: case OpenACCClauseKind::DeviceResident: case OpenACCClauseKind::Host: diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp index f263082cc13e9a..0b2a62a219358c 100644 --- a/clang/lib/Serialization/ASTWriter.cpp +++ b/clang/lib/Serialization/ASTWriter.cpp @@ -8356,6 +8356,12 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { writeOpenACCVarList(AC); return; } + case OpenACCClauseKind::Detach: { + const auto *DC = cast<OpenACCDetachClause>(C); + writeSourceLocation(DC->getLParenLoc()); + writeOpenACCVarList(DC); + return; + } case OpenACCClauseKind::DevicePtr: { const auto *DPC = cast<OpenACCDevicePtrClause>(C); writeSourceLocation(DPC->getLParenLoc()); @@ -8501,7 +8507,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { case OpenACCClauseKind::NoHost: case OpenACCClauseKind::UseDevice: case OpenACCClauseKind::Delete: - case OpenACCClauseKind::Detach: 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 bcc68a18232598..003dc03f342c15 100644 --- a/clang/test/AST/ast-print-openacc-data-construct.cpp +++ b/clang/test/AST/ast-print-openacc-data-construct.cpp @@ -117,4 +117,7 @@ void foo() { // CHECK: #pragma acc host_data if_present #pragma acc host_data use_device(i) if_present ; +// CHECK: #pragma acc exit data copyout(i) detach(iPtr, arrayPtr[0]) +#pragma acc exit data copyout(i) detach(iPtr, arrayPtr[0]) + } diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c index 8003dad03c1c0d..0498fe16e512cd 100644 --- a/clang/test/ParserOpenACC/parse-clauses.c +++ b/clang/test/ParserOpenACC/parse-clauses.c @@ -488,14 +488,13 @@ void VarListClauses() { #pragma acc serial attach(IsPointer), self for(int i = 0; i < 5;++i) {} - // expected-error@+2{{expected ','}} - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented, clause ignored}} -#pragma acc serial detach(s.array[s.value] s.array[s.value :5] ), self - for(int i = 0; i < 5;++i) {} + // expected-error@+4{{expected ','}} + // expected-error@+3{{expected pointer in 'detach' clause, type is 'char'}} + // expected-error@+2{{OpenACC sub-array is not allowed here}} + // expected-note@+1{{expected variable of pointer type}} +#pragma acc exit data copyout(s) detach(s.array[s.value] s.array[s.value :5]) - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented, clause ignored}} -#pragma acc serial detach(s.array[s.value : 5], s.value), self - for(int i = 0; i < 5;++i) {} +#pragma acc exit data copyout(s) detach(IsPointer) // expected-error@+1{{expected ','}} #pragma acc serial private(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 516658f8e01ff4..b1d682c0ea5b39 100644 --- a/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c +++ b/clang/test/SemaOpenACC/combined-construct-auto_seq_independent-clauses.c @@ -72,7 +72,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc parallel loop auto delete(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -189,7 +189,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc parallel loop delete(Var) auto for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -307,7 +307,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc parallel loop independent delete(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -424,7 +424,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc parallel loop delete(Var) independent for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -550,7 +550,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc parallel loop seq delete(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -673,7 +673,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc parallel loop delete(Var) seq for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' 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 1a4bd94063e6c5..cc8d8adbdc9f1c 100644 --- a/clang/test/SemaOpenACC/combined-construct-device_type-clause.c +++ b/clang/test/SemaOpenACC/combined-construct-device_type-clause.c @@ -99,8 +99,7 @@ void uses() { // expected-note@+1{{previous clause is here}} #pragma acc serial loop device_type(*) delete(Var) for(int i = 0; i < 5; ++i); - // expected-error@+2{{OpenACC clause 'detach' may not follow a 'device_type' clause in a 'kernels loop' construct}} - // expected-note@+1{{previous clause is here}} + // expected-error@+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@+2{{OpenACC clause 'device' 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 dff8edb7f3dbfd..6c7233e06d7758 100644 --- a/clang/test/SemaOpenACC/compute-construct-device_type-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-device_type-clause.c @@ -103,8 +103,7 @@ void uses() { // expected-note@+1{{previous clause is here}} #pragma acc kernels device_type(*) delete(Var) while(1); - // expected-error@+2{{OpenACC clause 'detach' may not follow a 'device_type' clause in a 'kernels' construct}} - // expected-note@+1{{previous clause is here}} + // expected-error@+1{{OpenACC 'detach' clause is not valid on 'kernels' directive}} #pragma acc kernels device_type(*) detach(Var) while(1); // expected-error@+2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'kernels' construct}} diff --git a/clang/test/SemaOpenACC/data-construct-detach-ast.cpp b/clang/test/SemaOpenACC/data-construct-detach-ast.cpp new file mode 100644 index 00000000000000..d5096cdc868d08 --- /dev/null +++ b/clang/test/SemaOpenACC/data-construct-detach-ast.cpp @@ -0,0 +1,61 @@ +// 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 exit data copyout(Global) detach(PointerParam) + // CHECK-NEXT: OpenACCExitDataConstruct{{.*}} exit data + // CHECK-NEXT: copyout clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Global' 'int' + // CHECK-NEXT: detach clause + // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}} 'PointerParam' 'float *' +} + +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 exit data copyout(Global) detach(t) + // CHECK-NEXT: OpenACCExitDataConstruct{{.*}} exit data + // CHECK-NEXT: copyout clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Global' 'int' + // CHECK-NEXT: detach clause + // CHECK-NEXT: DeclRefExpr{{.*}}'T *' lvalue ParmVar{{.*}} 't' 'T *' + + + // 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: OpenACCExitDataConstruct{{.*}} exit data + // CHECK-NEXT: copyout clause + // CHECK-NEXT: DeclRefExpr{{.*}}'Global' 'int' + // CHECK-NEXT: detach clause + // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 't' 'int *' + +} + +void Inst() { + int i; + TemplUses(&i); +} +#endif diff --git a/clang/test/SemaOpenACC/data-construct-detach-clause.c b/clang/test/SemaOpenACC/data-construct-detach-clause.c new file mode 100644 index 00000000000000..e75c95d99ec078 --- /dev/null +++ b/clang/test/SemaOpenACC/data-construct-detach-clause.c @@ -0,0 +1,68 @@ +// RUN: %clang_cc1 %s -fopenacc -verify + +struct S { + int IntMem; + int *PtrMem; +}; + +void uses() { + int LocalInt; + int *LocalPtr; + int Array[5]; + int *PtrArray[5]; + struct S s; + + // expected-error@+1{{expected pointer in 'detach' clause, type is 'int'}} +#pragma acc exit data copyout(LocalInt) detach(LocalInt) + ; + + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, member of a composite variable, or composite variable member}} +#pragma acc exit data copyout(LocalInt) detach(&LocalInt) + ; + + + // expected-error@+1{{expected pointer in 'detach' clause, type is 'int[5]'}} +#pragma acc exit data copyout(LocalInt) detach(Array) + + // expected-error@+1{{expected pointer in 'detach' clause, type is 'int'}} +#pragma acc exit data copyout(LocalInt) detach(Array[0]) + ; + + // expected-error@+2{{OpenACC sub-array is not allowed here}} + // expected-note@+1{{expected variable of pointer type}} +#pragma acc exit data copyout(LocalInt) detach(Array[0:1]) + ; + + // expected-error@+1{{expected pointer in 'detach' clause, type is 'int *[5]'}} +#pragma acc exit data copyout(LocalInt) detach(PtrArray) + ; + +#pragma acc exit data copyout(LocalInt) detach(PtrArray[0]) + ; + + // expected-error@+2{{OpenACC sub-array is not allowed here}} + // expected-note@+1{{expected variable of pointer type}} +#pragma acc exit data copyout(LocalInt) detach(PtrArray[0:1]) + ; + + // expected-error@+1{{expected pointer in 'detach' clause, type is 'struct S'}} +#pragma acc exit data copyout(LocalInt) detach(s) + ; + + // expected-error@+1{{expected pointer in 'detach' clause, type is 'int'}} +#pragma acc exit data copyout(LocalInt) detach(s.IntMem) + ; + +#pragma acc exit data copyout(LocalInt) detach(s.PtrMem) + ; + + // expected-error@+1{{OpenACC 'detach' clause is not valid on 'data' directive}} +#pragma acc data copyin(LocalInt) detach(PtrArray[0]) + ; + // 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.cpp b/clang/test/SemaOpenACC/data-construct.cpp index 0be88c7724e613..1fcf147b0f1b34 100644 --- a/clang/test/SemaOpenACC/data-construct.cpp +++ b/clang/test/SemaOpenACC/data-construct.cpp @@ -75,8 +75,7 @@ void AtLeastOneOf() { #pragma acc exit data copyout(Var) // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc exit data delete(Var) - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} -#pragma acc exit data detach(Var) +#pragma acc exit data detach(VarPtr) // OpenACC TODO: The following 'exit data' directives should diagnose, since // they don't have at least one of the above clauses. 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 a9ef1a03654a5a..e1e35d6a7b36e0 100644 --- a/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c +++ b/clang/test/SemaOpenACC/loop-construct-auto_seq_independent-clauses.c @@ -77,7 +77,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc loop auto delete(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -211,7 +211,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc loop delete(Var) auto for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -346,7 +346,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc loop independent delete(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -480,7 +480,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc loop delete(Var) independent for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -623,7 +623,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc loop seq delete(Var) for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' not yet implemented}} @@ -763,7 +763,7 @@ void uses() { // expected-warning@+1{{OpenACC clause 'delete' not yet implemented}} #pragma acc loop delete(Var) seq for(unsigned i = 0; i < 5; ++i); - // expected-warning@+1{{OpenACC clause 'detach' not yet implemented}} + // expected-error@+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@+1{{OpenACC clause 'device' 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 ad572ad75bca65..94eb2672e8111e 100644 --- a/clang/test/SemaOpenACC/loop-construct-device_type-clause.c +++ b/clang/test/SemaOpenACC/loop-construct-device_type-clause.c @@ -91,8 +91,7 @@ void uses() { // expected-note@+1{{previous clause is here}} #pragma acc loop device_type(*) delete(Var) for(int i = 0; i < 5; ++i); - // expected-error@+2{{OpenACC clause 'detach' may not follow a 'device_type' clause in a 'loop' construct}} - // expected-note@+1{{previous clause is here}} + // expected-error@+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@+2{{OpenACC clause 'device' may not follow a 'device_type' clause in a 'loop' construct}} diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp index c5fdb9065a1c7a..8d6994128f2f07 100644 --- a/clang/tools/libclang/CIndex.cpp +++ b/clang/tools/libclang/CIndex.cpp @@ -2886,6 +2886,11 @@ void OpenACCClauseEnqueue::VisitCreateClause(const OpenACCCreateClause &C) { void OpenACCClauseEnqueue::VisitAttachClause(const OpenACCAttachClause &C) { VisitVarList(C); } + +void OpenACCClauseEnqueue::VisitDetachClause(const OpenACCDetachClause &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
