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