Author: Erich Keane Date: 2024-04-30T11:28:37-07:00 New Revision: fa67986d5b309ddd4c2ea74e8a5eeb0559aa8022
URL: https://github.com/llvm/llvm-project/commit/fa67986d5b309ddd4c2ea74e8a5eeb0559aa8022 DIFF: https://github.com/llvm/llvm-project/commit/fa67986d5b309ddd4c2ea74e8a5eeb0559aa8022.diff LOG: [OpenACC] Private Clause on Compute Constructs (#90521) The private clause is the first that takes a 'var-list', thus this has a lot of additional work to enable the var-list type. A 'var' is a traditional variable reference, subscript, member-expression, or array-section, so checking of these is pretty minor. Note: This ran into some issues with array-sections (aka sub-arrays) that will be fixed in a follow-up patch. Added: clang/test/SemaOpenACC/compute-construct-private-clause.c clang/test/SemaOpenACC/compute-construct-private-clause.cpp clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp Modified: clang/include/clang/AST/OpenACCClause.h clang/include/clang/Basic/DiagnosticSemaKinds.td clang/include/clang/Basic/OpenACCClauses.def clang/include/clang/Parse/Parser.h clang/include/clang/Sema/SemaOpenACC.h clang/include/clang/Serialization/ASTRecordReader.h clang/include/clang/Serialization/ASTRecordWriter.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-compute-construct.cpp clang/test/ParserOpenACC/parse-cache-construct.c clang/test/ParserOpenACC/parse-clauses.c clang/tools/libclang/CIndex.cpp Removed: ################################################################################ diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h index 704f15ff60edb2..bb4cb1f5d5080b 100644 --- a/clang/include/clang/AST/OpenACCClause.h +++ b/clang/include/clang/AST/OpenACCClause.h @@ -156,51 +156,50 @@ class OpenACCSelfClause : public OpenACCClauseWithCondition { Expr *ConditionExpr, SourceLocation EndLoc); }; -/// Represents a clause that has one or more IntExprs. It does not own the -/// IntExprs, but provides 'children' and other accessors. -class OpenACCClauseWithIntExprs : public OpenACCClauseWithParams { - MutableArrayRef<Expr *> IntExprs; +/// Represents a clause that has one or more expressions associated with it. +class OpenACCClauseWithExprs : public OpenACCClauseWithParams { + MutableArrayRef<Expr *> Exprs; protected: - OpenACCClauseWithIntExprs(OpenACCClauseKind K, SourceLocation BeginLoc, - SourceLocation LParenLoc, SourceLocation EndLoc) + OpenACCClauseWithExprs(OpenACCClauseKind K, SourceLocation BeginLoc, + SourceLocation LParenLoc, SourceLocation EndLoc) : OpenACCClauseWithParams(K, BeginLoc, LParenLoc, EndLoc) {} /// Used only for initialization, the leaf class can initialize this to /// trailing storage. - void setIntExprs(MutableArrayRef<Expr *> NewIntExprs) { - assert(IntExprs.empty() && "Cannot change IntExprs list"); - IntExprs = NewIntExprs; + void setExprs(MutableArrayRef<Expr *> NewExprs) { + assert(Exprs.empty() && "Cannot change Exprs list"); + Exprs = NewExprs; } - /// Gets the entire list of integer expressions, but leave it to the + /// Gets the entire list of expressions, but leave it to the /// individual clauses to expose this how they'd like. - llvm::ArrayRef<Expr *> getIntExprs() const { return IntExprs; } + llvm::ArrayRef<Expr *> getExprs() const { return Exprs; } public: child_range children() { - return child_range(reinterpret_cast<Stmt **>(IntExprs.begin()), - reinterpret_cast<Stmt **>(IntExprs.end())); + return child_range(reinterpret_cast<Stmt **>(Exprs.begin()), + reinterpret_cast<Stmt **>(Exprs.end())); } const_child_range children() const { child_range Children = - const_cast<OpenACCClauseWithIntExprs *>(this)->children(); + const_cast<OpenACCClauseWithExprs *>(this)->children(); return const_child_range(Children.begin(), Children.end()); } }; class OpenACCNumGangsClause final - : public OpenACCClauseWithIntExprs, + : public OpenACCClauseWithExprs, public llvm::TrailingObjects<OpenACCNumGangsClause, Expr *> { OpenACCNumGangsClause(SourceLocation BeginLoc, SourceLocation LParenLoc, ArrayRef<Expr *> IntExprs, SourceLocation EndLoc) - : OpenACCClauseWithIntExprs(OpenACCClauseKind::NumGangs, BeginLoc, - LParenLoc, EndLoc) { + : OpenACCClauseWithExprs(OpenACCClauseKind::NumGangs, BeginLoc, LParenLoc, + EndLoc) { std::uninitialized_copy(IntExprs.begin(), IntExprs.end(), getTrailingObjects<Expr *>()); - setIntExprs(MutableArrayRef(getTrailingObjects<Expr *>(), IntExprs.size())); + setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), IntExprs.size())); } public: @@ -209,35 +208,35 @@ class OpenACCNumGangsClause final ArrayRef<Expr *> IntExprs, SourceLocation EndLoc); llvm::ArrayRef<Expr *> getIntExprs() { - return OpenACCClauseWithIntExprs::getIntExprs(); + return OpenACCClauseWithExprs::getExprs(); } llvm::ArrayRef<Expr *> getIntExprs() const { - return OpenACCClauseWithIntExprs::getIntExprs(); + return OpenACCClauseWithExprs::getExprs(); } }; /// Represents one of a handful of clauses that have a single integer /// expression. -class OpenACCClauseWithSingleIntExpr : public OpenACCClauseWithIntExprs { +class OpenACCClauseWithSingleIntExpr : public OpenACCClauseWithExprs { Expr *IntExpr; protected: OpenACCClauseWithSingleIntExpr(OpenACCClauseKind K, SourceLocation BeginLoc, SourceLocation LParenLoc, Expr *IntExpr, SourceLocation EndLoc) - : OpenACCClauseWithIntExprs(K, BeginLoc, LParenLoc, EndLoc), + : OpenACCClauseWithExprs(K, BeginLoc, LParenLoc, EndLoc), IntExpr(IntExpr) { - setIntExprs(MutableArrayRef<Expr *>{&this->IntExpr, 1}); + setExprs(MutableArrayRef<Expr *>{&this->IntExpr, 1}); } public: - bool hasIntExpr() const { return !getIntExprs().empty(); } + bool hasIntExpr() const { return !getExprs().empty(); } const Expr *getIntExpr() const { - return hasIntExpr() ? getIntExprs()[0] : nullptr; + return hasIntExpr() ? getExprs()[0] : nullptr; } - Expr *getIntExpr() { return hasIntExpr() ? getIntExprs()[0] : nullptr; }; + Expr *getIntExpr() { return hasIntExpr() ? getExprs()[0] : nullptr; }; }; class OpenACCNumWorkersClause : public OpenACCClauseWithSingleIntExpr { @@ -261,6 +260,40 @@ class OpenACCVectorLengthClause : public OpenACCClauseWithSingleIntExpr { Expr *IntExpr, SourceLocation EndLoc); }; +/// Represents a clause with one or more 'var' objects, represented as an expr, +/// as its arguments. Var-list is expected to be stored in trailing storage. +/// For now, we're just storing the original expression in its entirety, unlike +/// OMP which has to do a bunch of work to create a private. +class OpenACCClauseWithVarList : public OpenACCClauseWithExprs { +protected: + OpenACCClauseWithVarList(OpenACCClauseKind K, SourceLocation BeginLoc, + SourceLocation LParenLoc, SourceLocation EndLoc) + : OpenACCClauseWithExprs(K, BeginLoc, LParenLoc, EndLoc) {} + +public: + ArrayRef<Expr *> getVarList() { return getExprs(); } + ArrayRef<Expr *> getVarList() const { return getExprs(); } +}; + +class OpenACCPrivateClause final + : public OpenACCClauseWithVarList, + public llvm::TrailingObjects<OpenACCPrivateClause, Expr *> { + + OpenACCPrivateClause(SourceLocation BeginLoc, SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, SourceLocation EndLoc) + : OpenACCClauseWithVarList(OpenACCClauseKind::Private, BeginLoc, + LParenLoc, EndLoc) { + std::uninitialized_copy(VarList.begin(), VarList.end(), + getTrailingObjects<Expr *>()); + setExprs(MutableArrayRef(getTrailingObjects<Expr *>(), VarList.size())); + } + +public: + static OpenACCPrivateClause * + Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, SourceLocation EndLoc); +}; + template <class Impl> class OpenACCClauseVisitor { Impl &getDerived() { return static_cast<Impl &>(*this); } diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td index 5cb3c808957b30..4b074b853bfe65 100644 --- a/clang/include/clang/Basic/DiagnosticSemaKinds.td +++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td @@ -12307,4 +12307,7 @@ def err_acc_num_gangs_num_args "OpenACC 'num_gangs' " "%select{|clause: '%1' directive expects maximum of %2, %3 were " "provided}0">; +def err_acc_not_a_var_ref + : Error<"OpenACC variable is not a valid variable name, sub-array, array " + "element, or composite variable member">; } // end of sema component. diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def index dd5792e7ca8c39..6c3c2db66ef0cf 100644 --- a/clang/include/clang/Basic/OpenACCClauses.def +++ b/clang/include/clang/Basic/OpenACCClauses.def @@ -20,6 +20,7 @@ VISIT_CLAUSE(If) VISIT_CLAUSE(Self) VISIT_CLAUSE(NumGangs) VISIT_CLAUSE(NumWorkers) +VISIT_CLAUSE(Private) VISIT_CLAUSE(VectorLength) #undef VISIT_CLAUSE diff --git a/clang/include/clang/Parse/Parser.h b/clang/include/clang/Parse/Parser.h index 81aab8c888ab65..daefd4f28f011a 100644 --- a/clang/include/clang/Parse/Parser.h +++ b/clang/include/clang/Parse/Parser.h @@ -3654,11 +3654,12 @@ class Parser : public CodeCompletionHandler { ExprResult ParseOpenACCIDExpression(); /// Parses the variable list for the `cache` construct. void ParseOpenACCCacheVarList(); + + using OpenACCVarParseResult = std::pair<ExprResult, OpenACCParseCanContinue>; /// Parses a single variable in a variable list for OpenACC. - bool ParseOpenACCVar(); - /// Parses the variable list for the variety of clauses that take a var-list, - /// including the optional Special Token listed for some,based on clause type. - bool ParseOpenACCClauseVarList(OpenACCClauseKind Kind); + OpenACCVarParseResult ParseOpenACCVar(); + /// Parses the variable list for the variety of places that take a var-list. + llvm::SmallVector<Expr *> ParseOpenACCVarList(); /// Parses any parameters for an OpenACC Clause, including required/optional /// parens. OpenACCClauseParseResult diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h index da19503c2902fd..edb0cbb7c5d552 100644 --- a/clang/include/clang/Sema/SemaOpenACC.h +++ b/clang/include/clang/Sema/SemaOpenACC.h @@ -48,8 +48,12 @@ class SemaOpenACC : public SemaBase { SmallVector<Expr *> IntExprs; }; + struct VarListDetails { + SmallVector<Expr *> VarList; + }; + std::variant<std::monostate, DefaultDetails, ConditionDetails, - IntExprDetails> + IntExprDetails, VarListDetails> Details = std::monostate{}; public: @@ -112,6 +116,16 @@ class SemaOpenACC : public SemaBase { return const_cast<OpenACCParsedClause *>(this)->getIntExprs(); } + ArrayRef<Expr *> getVarList() { + assert(ClauseKind == OpenACCClauseKind::Private && + "Parsed clause kind does not have a var-list"); + return std::get<VarListDetails>(Details).VarList; + } + + ArrayRef<Expr *> getVarList() const { + return const_cast<OpenACCParsedClause *>(this)->getVarList(); + } + void setLParenLoc(SourceLocation EndLoc) { LParenLoc = EndLoc; } void setEndLoc(SourceLocation EndLoc) { ClauseRange.setEnd(EndLoc); } @@ -147,7 +161,19 @@ class SemaOpenACC : public SemaBase { ClauseKind == OpenACCClauseKind::NumWorkers || ClauseKind == OpenACCClauseKind::VectorLength) && "Parsed clause kind does not have a int exprs"); - Details = IntExprDetails{IntExprs}; + Details = IntExprDetails{std::move(IntExprs)}; + } + + void setVarListDetails(ArrayRef<Expr *> VarList) { + assert(ClauseKind == OpenACCClauseKind::Private && + "Parsed clause kind does not have a var-list"); + Details = VarListDetails{{VarList.begin(), VarList.end()}}; + } + + void setVarListDetails(llvm::SmallVector<Expr *> &&VarList) { + assert(ClauseKind == OpenACCClauseKind::Private && + "Parsed clause kind does not have a var-list"); + Details = VarListDetails{std::move(VarList)}; } }; @@ -194,6 +220,10 @@ class SemaOpenACC : public SemaBase { ExprResult ActOnIntExpr(OpenACCDirectiveKind DK, OpenACCClauseKind CK, SourceLocation Loc, Expr *IntExpr); + /// Called when encountering a 'var' for OpenACC, ensures it is actually a + /// declaration reference to a variable of the correct type. + ExprResult ActOnVar(Expr *VarExpr); + /// Checks and creates an Array Section used in an OpenACC construct/clause. ExprResult ActOnArraySectionExpr(Expr *Base, SourceLocation LBLoc, Expr *LowerBound, diff --git a/clang/include/clang/Serialization/ASTRecordReader.h b/clang/include/clang/Serialization/ASTRecordReader.h index 06b80f266a9441..1e11d2d5e42f95 100644 --- a/clang/include/clang/Serialization/ASTRecordReader.h +++ b/clang/include/clang/Serialization/ASTRecordReader.h @@ -269,6 +269,9 @@ class ASTRecordReader /// Read an OpenMP children, advancing Idx. void readOMPChildren(OMPChildren *Data); + /// Read a list of Exprs used for a var-list. + llvm::SmallVector<Expr *> readOpenACCVarList(); + /// Read an OpenACC clause, advancing Idx. OpenACCClause *readOpenACCClause(); diff --git a/clang/include/clang/Serialization/ASTRecordWriter.h b/clang/include/clang/Serialization/ASTRecordWriter.h index 1feb8fcbacf772..8b1da49bd4c576 100644 --- a/clang/include/clang/Serialization/ASTRecordWriter.h +++ b/clang/include/clang/Serialization/ASTRecordWriter.h @@ -15,6 +15,7 @@ #define LLVM_CLANG_SERIALIZATION_ASTRECORDWRITER_H #include "clang/AST/AbstractBasicWriter.h" +#include "clang/AST/OpenACCClause.h" #include "clang/AST/OpenMPClause.h" #include "clang/Serialization/ASTWriter.h" #include "clang/Serialization/SourceLocationEncoding.h" @@ -293,6 +294,8 @@ class ASTRecordWriter /// Writes data related to the OpenMP directives. void writeOMPChildren(OMPChildren *Data); + void writeOpenACCVarList(const OpenACCClauseWithVarList *C); + /// Writes out a single OpenACC Clause. void writeOpenACCClause(const OpenACCClause *C); diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp index b59578d80ed820..885f3b7618ec55 100644 --- a/clang/lib/AST/OpenACCClause.cpp +++ b/clang/lib/AST/OpenACCClause.cpp @@ -134,6 +134,16 @@ OpenACCNumGangsClause *OpenACCNumGangsClause::Create(const ASTContext &C, return new (Mem) OpenACCNumGangsClause(BeginLoc, LParenLoc, IntExprs, EndLoc); } +OpenACCPrivateClause *OpenACCPrivateClause::Create(const ASTContext &C, + SourceLocation BeginLoc, + SourceLocation LParenLoc, + ArrayRef<Expr *> VarList, + SourceLocation EndLoc) { + void *Mem = C.Allocate( + OpenACCPrivateClause::totalSizeToAlloc<Expr *>(VarList.size())); + return new (Mem) OpenACCPrivateClause(BeginLoc, LParenLoc, VarList, EndLoc); +} + //===----------------------------------------------------------------------===// // OpenACC clauses printing methods //===----------------------------------------------------------------------===// @@ -181,3 +191,10 @@ void OpenACCClausePrinter::VisitVectorLengthClause( printExpr(C.getIntExpr()); OS << ")"; } + +void OpenACCClausePrinter::VisitPrivateClause(const OpenACCPrivateClause &C) { + OS << "private("; + 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 a95f5c6103e24d..973f6f97bae0bf 100644 --- a/clang/lib/AST/StmtProfile.cpp +++ b/clang/lib/AST/StmtProfile.cpp @@ -2509,6 +2509,12 @@ void OpenACCClauseProfiler::VisitNumWorkersClause( Profiler.VisitStmt(Clause.getIntExpr()); } +void OpenACCClauseProfiler::VisitPrivateClause( + const OpenACCPrivateClause &Clause) { + for (auto *E : Clause.getVarList()) + Profiler.VisitStmt(E); +} + void OpenACCClauseProfiler::VisitVectorLengthClause( const OpenACCVectorLengthClause &Clause) { assert(Clause.hasIntExpr() && diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp index 8f0a9a9b0ed0bc..89f50d6dacfd23 100644 --- a/clang/lib/AST/TextNodeDumper.cpp +++ b/clang/lib/AST/TextNodeDumper.cpp @@ -401,6 +401,7 @@ void TextNodeDumper::Visit(const OpenACCClause *C) { case OpenACCClauseKind::Self: case OpenACCClauseKind::NumGangs: case OpenACCClauseKind::NumWorkers: + case OpenACCClauseKind::Private: case OpenACCClauseKind::VectorLength: // The condition expression will be printed as a part of the 'children', // but print 'clause' here so it is clear what is happening from the dump. diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp index 29326f5d993a9d..2d1ec6539b2fd1 100644 --- a/clang/lib/Parse/ParseOpenACC.cpp +++ b/clang/lib/Parse/ParseOpenACC.cpp @@ -86,6 +86,10 @@ OpenACCClauseKind getOpenACCClauseKind(Token Tok) { if (Tok.is(tok::kw_if)) return OpenACCClauseKind::If; + // 'private' is also a keyword, make sure we pare it correctly. + if (Tok.is(tok::kw_private)) + return OpenACCClauseKind::Private; + if (!Tok.is(tok::identifier)) return OpenACCClauseKind::Invalid; @@ -682,28 +686,6 @@ bool Parser::ParseOpenACCIntExprList(OpenACCDirectiveKind DK, return false; } -bool Parser::ParseOpenACCClauseVarList(OpenACCClauseKind Kind) { - // FIXME: Future clauses will require 'special word' parsing, check for one, - // then parse it based on whether it is a clause that requires a 'special - // word'. - (void)Kind; - - // If the var parsing fails, skip until the end of the directive as this is - // an expression and gets messy if we try to continue otherwise. - if (ParseOpenACCVar()) - return true; - - while (!getCurToken().isOneOf(tok::r_paren, tok::annot_pragma_openacc_end)) { - ExpectAndConsume(tok::comma); - - // If the var parsing fails, skip until the end of the directive as this is - // an expression and gets messy if we try to continue otherwise. - if (ParseOpenACCVar()) - return true; - } - return false; -} - /// OpenACC 3.3 Section 2.4: /// The argument to the device_type clause is a comma-separated list of one or /// more device architecture name identifiers, or an asterisk. @@ -917,28 +899,19 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams( case OpenACCClauseKind::CopyIn: tryParseAndConsumeSpecialTokenKind( *this, OpenACCSpecialTokenKind::ReadOnly, ClauseKind); - if (ParseOpenACCClauseVarList(ClauseKind)) { - Parens.skipToEnd(); - return OpenACCCanContinue(); - } + ParseOpenACCVarList(); break; case OpenACCClauseKind::Create: case OpenACCClauseKind::CopyOut: tryParseAndConsumeSpecialTokenKind(*this, OpenACCSpecialTokenKind::Zero, ClauseKind); - if (ParseOpenACCClauseVarList(ClauseKind)) { - Parens.skipToEnd(); - return OpenACCCanContinue(); - } + ParseOpenACCVarList(); break; case OpenACCClauseKind::Reduction: // If we're missing a clause-kind (or it is invalid), see if we can parse // the var-list anyway. ParseReductionOperator(*this); - if (ParseOpenACCClauseVarList(ClauseKind)) { - Parens.skipToEnd(); - return OpenACCCanContinue(); - } + ParseOpenACCVarList(); break; case OpenACCClauseKind::Self: // The 'self' clause is a var-list instead of a 'condition' in the case of @@ -958,12 +931,11 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams( case OpenACCClauseKind::Link: case OpenACCClauseKind::NoCreate: case OpenACCClauseKind::Present: - case OpenACCClauseKind::Private: case OpenACCClauseKind::UseDevice: - if (ParseOpenACCClauseVarList(ClauseKind)) { - Parens.skipToEnd(); - return OpenACCCanContinue(); - } + ParseOpenACCVarList(); + break; + case OpenACCClauseKind::Private: + ParsedClause.setVarListDetails(ParseOpenACCVarList()); break; case OpenACCClauseKind::Collapse: { tryParseAndConsumeSpecialTokenKind(*this, OpenACCSpecialTokenKind::Force, @@ -1227,16 +1199,51 @@ ExprResult Parser::ParseOpenACCBindClauseArgument() { /// OpenACC 3.3, section 1.6: /// In this spec, a 'var' (in italics) is one of the following: -/// - a variable name (a scalar, array, or compisite variable name) +/// - a variable name (a scalar, array, or composite variable name) /// - a subarray specification with subscript ranges /// - an array element /// - a member of a composite variable /// - a common block name between slashes (fortran only) -bool Parser::ParseOpenACCVar() { +Parser::OpenACCVarParseResult Parser::ParseOpenACCVar() { OpenACCArraySectionRAII ArraySections(*this); - ExprResult Res = - getActions().CorrectDelayedTyposInExpr(ParseAssignmentExpression()); - return Res.isInvalid(); + + ExprResult Res = ParseAssignmentExpression(); + if (!Res.isUsable()) + return {Res, OpenACCParseCanContinue::Cannot}; + + Res = getActions().CorrectDelayedTyposInExpr(Res.get()); + if (!Res.isUsable()) + return {Res, OpenACCParseCanContinue::Can}; + + Res = getActions().OpenACC().ActOnVar(Res.get()); + + return {Res, OpenACCParseCanContinue::Can}; +} + +llvm::SmallVector<Expr *> Parser::ParseOpenACCVarList() { + llvm::SmallVector<Expr *> Vars; + + auto [Res, CanContinue] = ParseOpenACCVar(); + if (Res.isUsable()) { + Vars.push_back(Res.get()); + } else if (CanContinue == OpenACCParseCanContinue::Cannot) { + SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end, StopBeforeMatch); + return Vars; + } + + while (!getCurToken().isOneOf(tok::r_paren, tok::annot_pragma_openacc_end)) { + ExpectAndConsume(tok::comma); + + auto [Res, CanContinue] = ParseOpenACCVar(); + + if (Res.isUsable()) { + Vars.push_back(Res.get()); + } else if (CanContinue == OpenACCParseCanContinue::Cannot) { + SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end, StopBeforeMatch); + return Vars; + } + } + return Vars; } /// OpenACC 3.3, section 2.10: @@ -1259,24 +1266,9 @@ void Parser::ParseOpenACCCacheVarList() { // Sema/AST generation. } - bool FirstArray = true; - while (!getCurToken().isOneOf(tok::r_paren, tok::annot_pragma_openacc_end)) { - if (!FirstArray) - ExpectAndConsume(tok::comma); - FirstArray = false; - - // OpenACC 3.3, section 2.10: - // A 'var' in a cache directive must be a single array element or a simple - // subarray. In C and C++, a simple subarray is an array name followed by - // an extended array range specification in brackets, with a start and - // length such as: - // - // arr[lower:length] - // - if (ParseOpenACCVar()) - SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end, tok::comma, - StopBeforeMatch); - } + // ParseOpenACCVarList should leave us before a r-paren, so no need to skip + // anything here. + ParseOpenACCVarList(); } Parser::OpenACCDirectiveParseInfo Parser::ParseOpenACCDirective() { diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp index d5cfe82a5d7098..3ea81e0497c203 100644 --- a/clang/lib/Sema/SemaOpenACC.cpp +++ b/clang/lib/Sema/SemaOpenACC.cpp @@ -103,6 +103,18 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind, default: return false; } + case OpenACCClauseKind::Private: + switch (DirectiveKind) { + case OpenACCDirectiveKind::Parallel: + case OpenACCDirectiveKind::Serial: + case OpenACCDirectiveKind::Loop: + case OpenACCDirectiveKind::ParallelLoop: + case OpenACCDirectiveKind::SerialLoop: + case OpenACCDirectiveKind::KernelsLoop: + return true; + default: + return false; + } default: // Do nothing so we can go to the 'unimplemented' diagnostic instead. return true; @@ -303,6 +315,21 @@ SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses, getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(), Clause.getIntExprs()[0], Clause.getEndLoc()); } + case OpenACCClauseKind::Private: { + // Restrictions only properly implemented on 'compute' constructs, and + // 'compute' constructs are the only construct that can do anything with + // this yet, so skip/treat as unimplemented in this case. + if (!isOpenACCComputeDirectiveKind(Clause.getDirectiveKind())) + break; + + // ActOnVar ensured that everything is a valid variable reference, so there + // really isn't anything to do here. GCC does some duplicate-finding, though + // it isn't apparent in the standard where this is justified. + + return OpenACCPrivateClause::Create( + getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(), + Clause.getVarList(), Clause.getEndLoc()); + } default: break; } @@ -423,6 +450,52 @@ ExprResult SemaOpenACC::ActOnIntExpr(OpenACCDirectiveKind DK, return IntExpr; } +ExprResult SemaOpenACC::ActOnVar(Expr *VarExpr) { + // We still need to retain the array subscript/subarray exprs, so work on a + // copy. + Expr *CurVarExpr = VarExpr->IgnoreParenImpCasts(); + + // 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)) { + if (auto *SubScrpt = dyn_cast<ArraySubscriptExpr>(CurVarExpr)) + CurVarExpr = SubScrpt->getBase()->IgnoreParenImpCasts(); + else + CurVarExpr = + cast<ArraySectionExpr>(CurVarExpr)->getBase()->IgnoreParenImpCasts(); + } + + // References to a VarDecl are fine. + if (const auto *DRE = dyn_cast<DeclRefExpr>(CurVarExpr)) { + if (isa<VarDecl, NonTypeTemplateParmDecl>( + DRE->getDecl()->getCanonicalDecl())) + return VarExpr; + } + + // A MemberExpr that references a Field is valid. + 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)) + return VarExpr; + + // Nothing really we can do here, as these are dependent. So just return they + // are valid. + if (isa<DependentScopeDeclRefExpr, CXXDependentScopeMemberExpr>(CurVarExpr)) + return VarExpr; + + // There isn't really anything we can do in the case of a recovery expr, so + // skip the diagnostic rather than produce a confusing diagnostic. + if (isa<RecoveryExpr>(CurVarExpr)) + return ExprError(); + + Diag(VarExpr->getExprLoc(), diag::err_acc_not_a_var_ref); + return ExprError(); +} + ExprResult SemaOpenACC::ActOnArraySectionExpr(Expr *Base, SourceLocation LBLoc, Expr *LowerBound, SourceLocation ColonLoc, diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h index 28d3d1b79a7424..ababb5d02b5fdb 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -11104,13 +11104,15 @@ template <typename Derived> class OpenACCClauseTransform final : public OpenACCClauseVisitor<OpenACCClauseTransform<Derived>> { TreeTransform<Derived> &Self; + ArrayRef<const OpenACCClause *> ExistingClauses; SemaOpenACC::OpenACCParsedClause &ParsedClause; OpenACCClause *NewClause = nullptr; public: OpenACCClauseTransform(TreeTransform<Derived> &Self, + ArrayRef<const OpenACCClause *> ExistingClauses, SemaOpenACC::OpenACCParsedClause &PC) - : Self(Self), ParsedClause(PC) {} + : Self(Self), ExistingClauses(ExistingClauses), ParsedClause(PC) {} OpenACCClause *CreatedClause() const { return NewClause; } @@ -11196,6 +11198,31 @@ void OpenACCClauseTransform<Derived>::VisitNumGangsClause( ParsedClause.getLParenLoc(), ParsedClause.getIntExprs(), ParsedClause.getEndLoc()); } + +template <typename Derived> +void OpenACCClauseTransform<Derived>::VisitPrivateClause( + const OpenACCPrivateClause &C) { + llvm::SmallVector<Expr *> InstantiatedVarList; + + for (Expr *CurVar : C.getVarList()) { + ExprResult Res = Self.TransformExpr(CurVar); + + if (!Res.isUsable()) + return; + + Res = Self.getSema().OpenACC().ActOnVar(Res.get()); + + if (Res.isUsable()) + InstantiatedVarList.push_back(Res.get()); + } + ParsedClause.setVarListDetails(std::move(InstantiatedVarList)); + + NewClause = OpenACCPrivateClause::Create( + Self.getSema().getASTContext(), ParsedClause.getBeginLoc(), + ParsedClause.getLParenLoc(), ParsedClause.getVarList(), + ParsedClause.getEndLoc()); +} + template <typename Derived> void OpenACCClauseTransform<Derived>::VisitNumWorkersClause( const OpenACCNumWorkersClause &C) { @@ -11254,7 +11281,8 @@ OpenACCClause *TreeTransform<Derived>::TransformOpenACCClause( if (const auto *WithParms = dyn_cast<OpenACCClauseWithParams>(OldClause)) ParsedClause.setLParenLoc(WithParms->getLParenLoc()); - OpenACCClauseTransform<Derived> Transform{*this, ParsedClause}; + OpenACCClauseTransform<Derived> Transform{*this, ExistingClauses, + ParsedClause}; Transform.Visit(OldClause); return Transform.CreatedClause(); diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp index 0ef57a3ea804ef..143fbc7feb3ab7 100644 --- a/clang/lib/Serialization/ASTReader.cpp +++ b/clang/lib/Serialization/ASTReader.cpp @@ -11768,6 +11768,14 @@ void ASTRecordReader::readOMPChildren(OMPChildren *Data) { Data->getChildren()[I] = readStmt(); } +SmallVector<Expr *> ASTRecordReader::readOpenACCVarList() { + unsigned NumVars = readInt(); + llvm::SmallVector<Expr *> VarList; + for (unsigned I = 0; I < NumVars; ++I) + VarList.push_back(readSubExpr()); + return VarList; +} + OpenACCClause *ASTRecordReader::readOpenACCClause() { OpenACCClauseKind ClauseKind = readEnum<OpenACCClauseKind>(); SourceLocation BeginLoc = readSourceLocation(); @@ -11813,6 +11821,12 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { return OpenACCVectorLengthClause::Create(getContext(), BeginLoc, LParenLoc, IntExpr, EndLoc); } + case OpenACCClauseKind::Private: { + SourceLocation LParenLoc = readSourceLocation(); + llvm::SmallVector<Expr *> VarList = readOpenACCVarList(); + return OpenACCPrivateClause::Create(getContext(), BeginLoc, LParenLoc, + VarList, EndLoc); + } case OpenACCClauseKind::Finalize: case OpenACCClauseKind::IfPresent: case OpenACCClauseKind::Seq: @@ -11834,7 +11848,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { case OpenACCClauseKind::Link: case OpenACCClauseKind::NoCreate: case OpenACCClauseKind::Present: - case OpenACCClauseKind::Private: case OpenACCClauseKind::CopyOut: case OpenACCClauseKind::CopyIn: case OpenACCClauseKind::Create: diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp index c638c767216d94..b80b4ff14bd3c5 100644 --- a/clang/lib/Serialization/ASTWriter.cpp +++ b/clang/lib/Serialization/ASTWriter.cpp @@ -7703,6 +7703,12 @@ void ASTRecordWriter::writeOMPChildren(OMPChildren *Data) { AddStmt(Data->getChildren()[I]); } +void ASTRecordWriter::writeOpenACCVarList(const OpenACCClauseWithVarList *C) { + writeUInt32(C->getVarList().size()); + for (Expr *E : C->getVarList()) + AddStmt(E); +} + void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { writeEnum(C->getClauseKind()); writeSourceLocation(C->getBeginLoc()); @@ -7749,6 +7755,12 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { AddStmt(const_cast<Expr *>(NWC->getIntExpr())); return; } + case OpenACCClauseKind::Private: { + const auto *PC = cast<OpenACCPrivateClause>(C); + writeSourceLocation(PC->getLParenLoc()); + writeOpenACCVarList(PC); + return; + } case OpenACCClauseKind::Finalize: case OpenACCClauseKind::IfPresent: case OpenACCClauseKind::Seq: @@ -7770,7 +7782,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { case OpenACCClauseKind::Link: case OpenACCClauseKind::NoCreate: case OpenACCClauseKind::Present: - case OpenACCClauseKind::Private: case OpenACCClauseKind::CopyOut: case OpenACCClauseKind::CopyIn: case OpenACCClauseKind::Create: diff --git a/clang/test/AST/ast-print-openacc-compute-construct.cpp b/clang/test/AST/ast-print-openacc-compute-construct.cpp index a0e4294426a7fa..cd39ea087b3cad 100644 --- a/clang/test/AST/ast-print-openacc-compute-construct.cpp +++ b/clang/test/AST/ast-print-openacc-compute-construct.cpp @@ -35,5 +35,8 @@ void foo() { #pragma acc parallel vector_length((int)array[1]) while(true); +// CHECK: #pragma acc parallel private(i, array[1], array, array[1:2]) +#pragma acc parallel private(i, array[1], array, array[1:2]) + while(true); } diff --git a/clang/test/ParserOpenACC/parse-cache-construct.c b/clang/test/ParserOpenACC/parse-cache-construct.c index fd161c03c09f75..de26fc2b277a6b 100644 --- a/clang/test/ParserOpenACC/parse-cache-construct.c +++ b/clang/test/ParserOpenACC/parse-cache-construct.c @@ -25,11 +25,13 @@ void func() { } for (int i = 0; i < 10; ++i) { + // expected-error@+2{{expected expression}} // expected-warning@+1{{OpenACC construct 'cache' not yet implemented, pragma ignored}} #pragma acc cache() } for (int i = 0; i < 10; ++i) { + // expected-error@+3{{expected expression}} // expected-error@+2{{invalid OpenACC clause 'clause'}} // expected-warning@+1{{OpenACC construct 'cache' not yet implemented, pragma ignored}} #pragma acc cache() clause-list diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c index ee2cb2d1501dea..8a439a5ccd4bdc 100644 --- a/clang/test/ParserOpenACC/parse-clauses.c +++ b/clang/test/ParserOpenACC/parse-clauses.c @@ -405,7 +405,10 @@ void SelfUpdate() { #pragma acc update self for(;;){} - // expected-error@+3{{use of undeclared identifier 'zero'}} + // expected-error@+6{{use of undeclared identifier 'zero'}} + // expected-error@+5{{expected ','}} + // expected-error@+4{{expected expression}} + // expected-warning@+3{{OpenACC clause 'self' not yet implemented, clause ignored}} // expected-warning@+2{{OpenACC clause 'seq' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC construct 'update' not yet implemented, pragma ignored}} #pragma acc update self(zero : s.array[s.value : 5], s.value), seq @@ -450,11 +453,13 @@ void VarListClauses() { #pragma acc serial copy(, seq for(;;){} - // expected-error@+1{{expected expression}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'copy' not yet implemented, clause ignored}} #pragma acc serial copy() for(;;){} - // expected-error@+2{{expected expression}} + // expected-error@+3{{expected expression}} + // expected-warning@+2{{OpenACC clause 'copy' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copy(), seq for(;;){} @@ -494,24 +499,28 @@ void VarListClauses() { #pragma acc serial copy(HasMem.MemArr[1:3].array[1:2]), seq for(;;){} - // expected-error@+2{{expected expression}} + // expected-error@+3{{expected expression}} + // expected-warning@+2{{OpenACC clause 'copy' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copy(HasMem.MemArr[:]), seq for(;;){} - // expected-error@+2{{expected expression}} + // expected-error@+3{{expected expression}} + // expected-warning@+2{{OpenACC clause 'copy' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copy(HasMem.MemArr[::]), seq for(;;){} - // expected-error@+4{{expected expression}} - // expected-error@+3{{expected ']'}} - // expected-note@+2{{to match this '['}} + // expected-error@+5{{expected expression}} + // expected-error@+4{{expected ']'}} + // expected-note@+3{{to match this '['}} + // expected-warning@+2{{OpenACC clause 'copy' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copy(HasMem.MemArr[: :]), seq for(;;){} - // expected-error@+2{{expected expression}} + // expected-error@+3{{expected expression}} + // expected-warning@+2{{OpenACC clause 'copy' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copy(HasMem.MemArr[3:]), seq for(;;){} @@ -582,13 +591,11 @@ void VarListClauses() { #pragma acc serial detach(s.array[s.value : 5], s.value), seq for(;;){} - // expected-error@+3{{expected ','}} - // expected-warning@+2{{OpenACC clause 'private' not yet implemented, clause ignored}} + // expected-error@+2{{expected ','}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial private(s.array[s.value] s.array[s.value :5] ), seq for(;;){} - // expected-warning@+2{{OpenACC clause 'private' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial private(s.array[s.value : 5], s.value), seq for(;;){} @@ -691,7 +698,9 @@ void VarListClauses() { #pragma acc serial copyout(zero : s.array[s.value : 5], s.value), seq for(;;){} - // expected-error@+2{{use of undeclared identifier 'zero'}} + // expected-error@+4{{use of undeclared identifier 'zero'}} + // expected-error@+3{{expected ','}} + // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copyout(zero s.array[s.value : 5], s.value), seq for(;;){} @@ -714,7 +723,9 @@ void VarListClauses() { #pragma acc serial copyout(invalid:s.array[s.value : 5], s.value), seq for(;;){} - // expected-error@+2{{use of undeclared identifier 'invalid'}} + // expected-error@+4{{use of undeclared identifier 'invalid'}} + // expected-error@+3{{expected ','}} + // expected-warning@+2{{OpenACC clause 'copyout' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copyout(invalid s.array[s.value : 5], s.value), seq for(;;){} @@ -740,7 +751,9 @@ void VarListClauses() { #pragma acc serial create(zero : s.array[s.value : 5], s.value), seq for(;;){} - // expected-error@+2{{use of undeclared identifier 'zero'}} + // expected-error@+4{{use of undeclared identifier 'zero'}} + // expected-error@+3{{expected ','}} + // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial create(zero s.array[s.value : 5], s.value), seq for(;;){} @@ -763,7 +776,9 @@ void VarListClauses() { #pragma acc serial create(invalid:s.array[s.value : 5], s.value), seq for(;;){} - // expected-error@+2{{use of undeclared identifier 'invalid'}} + // expected-error@+4{{use of undeclared identifier 'invalid'}} + // expected-error@+3{{expected ','}} + // expected-warning@+2{{OpenACC clause 'create' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial create(invalid s.array[s.value : 5], s.value), seq for(;;){} @@ -789,7 +804,9 @@ void VarListClauses() { #pragma acc serial copyin(readonly : s.array[s.value : 5], s.value), seq for(;;){} - // expected-error@+2{{use of undeclared identifier 'readonly'}} + // expected-error@+4{{use of undeclared identifier 'readonly'}} + // expected-error@+3{{expected ','}} + // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copyin(readonly s.array[s.value : 5], s.value), seq for(;;){} @@ -812,7 +829,9 @@ void VarListClauses() { #pragma acc serial copyin(invalid:s.array[s.value : 5], s.value), seq for(;;){} - // expected-error@+2{{use of undeclared identifier 'invalid'}} + // expected-error@+4{{use of undeclared identifier 'invalid'}} + // expected-error@+3{{expected ','}} + // expected-warning@+2{{OpenACC clause 'copyin' not yet implemented, clause ignored}} // expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause ignored}} #pragma acc serial copyin(invalid s.array[s.value : 5], s.value), seq for(;;){} @@ -823,8 +842,9 @@ void ReductionClauseParsing() { // expected-error@+1{{expected '('}} #pragma acc serial reduction for(;;){} - // expected-error@+2{{missing reduction operator, expected '+', '*', 'max', 'min', '&', '|', '^', '&&', or '||', follwed by a ':'}} - // expected-error@+1{{expected expression}} + // expected-error@+3{{missing reduction operator, expected '+', '*', 'max', 'min', '&', '|', '^', '&&', or '||', follwed by a ':'}} + // expected-error@+2{{expected expression}} + // expected-warning@+1{{OpenACC clause 'reduction' not yet implemented, clause ignored}} #pragma acc serial reduction() for(;;){} // expected-error@+2{{missing reduction operator, expected '+', '*', 'max', 'min', '&', '|', '^', '&&', or '||', follwed by a ':'}} diff --git a/clang/test/SemaOpenACC/compute-construct-private-clause.c b/clang/test/SemaOpenACC/compute-construct-private-clause.c new file mode 100644 index 00000000000000..15775279fc8690 --- /dev/null +++ b/clang/test/SemaOpenACC/compute-construct-private-clause.c @@ -0,0 +1,145 @@ +// RUN: %clang_cc1 %s -fopenacc -verify + +struct Incomplete; +enum SomeE{ A }; +typedef struct IsComplete { + struct S { int A; } CompositeMember; + int ScalarMember; + float ArrayMember[5]; + enum SomeE EnumMember; + void *PointerMember; +} Complete; + +int GlobalInt; +float GlobalArray[5]; +void *GlobalPointer; +Complete GlobalComposite; + +void uses(int IntParam, void *PointerParam, float ArrayParam[5], Complete CompositeParam) { + int LocalInt; + void *LocalPointer; + float LocalArray[5]; + Complete LocalComposite; + + // Check Appertainment: +#pragma acc parallel private(LocalInt) + while(1); +#pragma acc serial private(LocalInt) + while(1); + // expected-error@+1{{OpenACC 'private' clause is not valid on 'kernels' directive}} +#pragma acc kernels private(LocalInt) + while(1); + + // Valid cases: +#pragma acc parallel private(LocalInt, LocalPointer, LocalArray) + while(1); +#pragma acc parallel private(LocalArray) + while(1); + // TODO OpenACC: Fix array sections, this should be allowed. + // expected-error@+1{{expected expression}} +#pragma acc parallel private(LocalArray[:]) + while(1); +#pragma acc parallel private(LocalArray[:5]) + while(1); + // TODO OpenACC: Fix array sections, this should be allowed. + // expected-error@+1{{expected expression}} +#pragma acc parallel private(LocalArray[2:]) + while(1); +#pragma acc parallel private(LocalArray[2:5]) + while(1); +#pragma acc parallel private(LocalArray[2]) + while(1); +#pragma acc parallel private(LocalComposite) + while(1); +#pragma acc parallel private(LocalComposite.EnumMember) + while(1); +#pragma acc parallel private(LocalComposite.ScalarMember) + while(1); +#pragma acc parallel private(LocalComposite.ArrayMember) + while(1); +#pragma acc parallel private(LocalComposite.ArrayMember[5]) + while(1); +#pragma acc parallel private(LocalComposite.PointerMember) + while(1); +#pragma acc parallel private(GlobalInt, GlobalArray, GlobalPointer, GlobalComposite) + while(1); +#pragma acc parallel private(GlobalArray[2], GlobalPointer[2], GlobalComposite.CompositeMember.A) + while(1); +#pragma acc parallel private(LocalComposite, GlobalComposite) + while(1); +#pragma acc parallel private(IntParam, PointerParam, ArrayParam, CompositeParam) + while(1); +#pragma acc parallel private(PointerParam[IntParam], ArrayParam[IntParam], CompositeParam.CompositeMember.A) + while(1); + +#pragma acc parallel private(LocalArray) private(LocalArray[2]) + while(1); + +#pragma acc parallel private(LocalArray, LocalArray[2]) + while(1); + +#pragma acc parallel private(LocalComposite, LocalComposite.ScalarMember) + while(1); + +#pragma acc parallel private(LocalComposite.CompositeMember.A, LocalComposite.ScalarMember) + while(1); + +#pragma acc parallel private(LocalComposite.CompositeMember.A) private(LocalComposite.ScalarMember) + while(1); + + Complete LocalComposite2; +#pragma acc parallel private(LocalComposite2.ScalarMember, LocalComposite2.ScalarMember) + while(1); + + // Invalid cases, arbitrary expressions. + struct Incomplete *I; + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(*I) + while(1); + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(GlobalInt + IntParam) + while(1); + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(+GlobalInt) + while(1); + + // TODO OpenACC: Fix array sections, this should be allowed. + // expected-error@+1{{expected expression}} +#pragma acc parallel private(PointerParam[:]) + while(1); +#pragma acc parallel private(PointerParam[:5]) + while(1); +#pragma acc parallel private(PointerParam[:IntParam]) + while(1); + // TODO OpenACC: Fix array sections, this should be allowed. + // expected-error@+1{{expected expression}} +#pragma acc parallel private(PointerParam[2:]) + while(1); +#pragma acc parallel private(PointerParam[2:5]) + while(1); +#pragma acc parallel private(PointerParam[2]) + while(1); + // TODO OpenACC: Fix array sections, this should be allowed. + // expected-error@+1{{expected expression}} +#pragma acc parallel private(ArrayParam[:]) + while(1); +#pragma acc parallel private(ArrayParam[:5]) + while(1); +#pragma acc parallel private(ArrayParam[:IntParam]) + while(1); + // TODO OpenACC: Fix array sections, this should be allowed. + // expected-error@+1{{expected expression}} +#pragma acc parallel private(ArrayParam[2:]) + while(1); +#pragma acc parallel private(ArrayParam[2:5]) + while(1); +#pragma acc parallel private(ArrayParam[2]) + while(1); + + // expected-error@+1{{OpenACC sub-array is not allowed here}} +#pragma acc parallel private((float*)ArrayParam[2:5]) + while(1); + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private((float)ArrayParam[2]) + while(1); +} diff --git a/clang/test/SemaOpenACC/compute-construct-private-clause.cpp b/clang/test/SemaOpenACC/compute-construct-private-clause.cpp new file mode 100644 index 00000000000000..4dd4e0d8029d67 --- /dev/null +++ b/clang/test/SemaOpenACC/compute-construct-private-clause.cpp @@ -0,0 +1,161 @@ +// RUN: %clang_cc1 %s -fopenacc -verify + +struct Incomplete; +enum SomeE{}; +typedef struct IsComplete { + struct S { int A; } CompositeMember; + int ScalarMember; + float ArrayMember[5]; + SomeE EnumMember; + char *PointerMember; +} Complete; + +int GlobalInt; +float GlobalArray[5]; +char *GlobalPointer; +Complete GlobalComposite; + +void uses(int IntParam, char *PointerParam, float ArrayParam[5], Complete CompositeParam, int &IntParamRef) { + int LocalInt; + char *LocalPointer; + float LocalArray[5]; + Complete LocalComposite; + + // Check Appertainment: + +#pragma acc parallel private(LocalInt) + while(true); +#pragma acc serial private(LocalInt) + while(true); + // expected-error@+1{{OpenACC 'private' clause is not valid on 'kernels' directive}} +#pragma acc kernels private(LocalInt) + while(true); + + // Valid cases: +#pragma acc parallel private(LocalInt, LocalPointer, LocalArray) + while(true); +#pragma acc parallel private(LocalArray) + while(true); +#pragma acc parallel private(LocalArray[2]) + while(true); +#pragma acc parallel private(LocalComposite) + while(true); +#pragma acc parallel private(LocalComposite.EnumMember) + while(true); +#pragma acc parallel private(LocalComposite.ScalarMember) + while(true); +#pragma acc parallel private(LocalComposite.ArrayMember) + while(true); +#pragma acc parallel private(LocalComposite.ArrayMember[5]) + while(true); +#pragma acc parallel private(LocalComposite.PointerMember) + while(true); +#pragma acc parallel private(GlobalInt, GlobalArray, GlobalPointer, GlobalComposite) + while(true); +#pragma acc parallel private(GlobalArray[2], GlobalPointer[2], GlobalComposite.CompositeMember.A) + while(true); +#pragma acc parallel private(LocalComposite, GlobalComposite) + while(true); +#pragma acc parallel private(IntParam, PointerParam, ArrayParam, CompositeParam) private(IntParamRef) + while(true); +#pragma acc parallel private(PointerParam[IntParam], ArrayParam[IntParam], CompositeParam.CompositeMember.A) + while(true); + + + // Invalid cases, arbitrary expressions. + Incomplete *I; + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(*I) + while(true); + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(GlobalInt + IntParam) + while(true); + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(+GlobalInt) + while(true); +} + +template<typename T, unsigned I, typename V> +void TemplUses(T t, T (&arrayT)[I], V TemplComp) { + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(+t) + while(true); + + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(+I) + while(true); + + // NTTP's are only valid if it is a reference to something. + // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} + // expected-note@#TEMPL_USES_INST{{in instantiation of}} +#pragma acc parallel private(I) + while(true); + + // expected-error@+1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} +#pragma acc parallel private(t, I) + while(true); + +#pragma acc parallel private(arrayT) + while(true); + +#pragma acc parallel private(TemplComp) + while(true); + +#pragma acc parallel private(TemplComp.PointerMember[5]) + while(true); + +#pragma acc parallel private(TemplComp.PointerMember[5]) private(TemplComp) + while(true); + + int *Pointer; +#pragma acc parallel private(Pointer[:I]) + while(true); +#pragma acc parallel private(Pointer[:t]) + while(true); + // TODO OpenACC: When fixing sub-arrays, this should be permitted}} + // expected-error@+1{{expected expression}} +#pragma acc parallel private(Pointer[1:]) + while(true); +} + +template<unsigned I, auto &NTTP_REF> +void NTTP() { + // NTTP's are only valid if it is a reference to something. + // expected-error@+2{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}} + // expected-note@#NTTP_INST{{in instantiation of}} +#pragma acc parallel private(I) + while(true); + +#pragma acc parallel private(NTTP_REF) + while(true); +} + +struct S { + int ThisMember; + int ThisMemberArray[5]; + + void foo(); +}; + +void S::foo() { +#pragma acc parallel private(ThisMember, this->ThisMemberArray[1]) + while(true); + +#pragma acc parallel private(ThisMemberArray[1:2]) + while(true); + +#pragma acc parallel private(this) + while(true); + +#pragma acc parallel private(ThisMember, this->ThisMember) + while(true); +} + +void Inst() { + static constexpr int NTTP_REFed = 1; + int i; + int Arr[5]; + Complete C; + TemplUses(i, Arr, C); // #TEMPL_USES_INST + NTTP<5, NTTP_REFed>(); // #NTTP_INST +} diff --git a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp new file mode 100644 index 00000000000000..5d10106724b907 --- /dev/null +++ b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp @@ -0,0 +1,552 @@ +// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s + +int Global; +short GlobalArray[5]; + +void NormalUses(float *PointerParam) { + // CHECK: FunctionDecl{{.*}}NormalUses + // CHECK: ParmVarDecl + // CHECK-NEXT: CompoundStmt + +#pragma acc parallel private(Global, GlobalArray[2]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' + // CHECK-NEXT: ArraySubscriptExpr{{.*}}'short' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'short *' <ArrayToPointerDecay> + // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]' + // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 2 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(GlobalArray, PointerParam[Global]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]' + // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *' + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(GlobalArray) private(PointerParam[Global]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]' + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySubscriptExpr{{.*}}'float' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'float *' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *' + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(GlobalArray, PointerParam[Global : Global]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]' + // CHECK-NEXT: ArraySectionExpr + // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}} 'PointerParam' 'float *' + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt +} + +// This example is an error typically, but we want to make sure we're properly +// capturing NTTPs until instantiation time. +template<unsigned I> +void UnInstTempl() { + // CHECK-NEXT: FunctionTemplateDecl{{.*}} UnInstTempl + // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}}referenced 'unsigned int' depth 0 index 0 I + // CHECK-NEXT: FunctionDecl{{.*}} UnInstTempl 'void ()' + // CHECK-NEXT: CompoundStmt +#pragma acc parallel private(I) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'unsigned int' NonTypeTemplateParm{{.*}}'I' 'unsigned int' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt +} + +template<auto &NTTP, typename T, typename U> +void TemplUses(T t, U u, T*PointerParam) { + // CHECK-NEXT: FunctionTemplateDecl + // CHECK-NEXT: NonTypeTemplateParmDecl {{.*}}referenced 'auto &' depth 0 index 0 NTTP + // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 1 T + // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 2 U + // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void (T, U, T *)' + // CHECK-NEXT: ParmVarDecl{{.*}} referenced t 'T' + // CHECK-NEXT: ParmVarDecl{{.*}} referenced u 'U' + // CHECK-NEXT: ParmVarDecl{{.*}} referenced PointerParam 'T *' + // CHECK-NEXT: CompoundStmt + + +#pragma acc parallel private(GlobalArray, PointerParam[Global]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]' + // CHECK-NEXT: ArraySubscriptExpr{{.*}}'T' lvalue + // CHECK-NEXT: DeclRefExpr{{.*}}'T *' lvalue ParmVar{{.*}}'PointerParam' 'T *' + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(t, u) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T' + // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(t) private(u) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T' + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(t) private(NTTP, u) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T' + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &' + // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(u[0]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySubscriptExpr{{.*}}'<dependent type>' lvalue + // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' + // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(u[0:t]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr + // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' + // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0 + // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl{{.*}}EndMarker + int EndMarker; + + // Check the instantiated versions of the above. + // CHECK-NEXT: FunctionDecl{{.*}} used TemplUses 'void (int, int *, int *)' implicit_instantiation + // CHECK-NEXT: TemplateArgument decl + // CHECK-NEXT: Var{{.*}} 'CEVar' 'const unsigned int' + // CHECK-NEXT: TemplateArgument type 'int' + // CHECK-NEXT: BuiltinType{{.*}} 'int' + // CHECK-NEXT: TemplateArgument type 'int[1]' + // CHECK-NEXT: ConstantArrayType{{.*}} 'int[1]' 1 + // CHECK-NEXT: BuiltinType{{.*}} 'int' + // CHECK-NEXT: ParmVarDecl{{.*}} used t 'int' + // CHECK-NEXT: ParmVarDecl{{.*}} used u 'int *' + // CHECK-NEXT: ParmVarDecl{{.*}} used PointerParam 'int *' + // CHECK-NEXT: CompoundStmt + +// #pragma acc parallel private(GlobalArray, PointerParam[Global]) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]' + // CHECK-NEXT: ArraySubscriptExpr{{.*}}'int' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}}'PointerParam' 'int *' + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +// #pragma acc parallel private(t, u) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int' + // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +// #pragma acc parallel private(t) private(u) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int' + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +// #pragma acc parallel private(t) private(NTTP, u) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int' + // CHECK-NEXT: private clause + // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue + // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP + // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int' + // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +// #pragma acc parallel private(u[0]) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySubscriptExpr{{.*}}'int' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' + // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +// #pragma acc parallel private(u[0:t]) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr + // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' + // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0 + // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + + // CHECK-NEXT: DeclStmt + // CHECK-NEXT: VarDecl{{.*}}EndMarker +} + +struct S { + // CHECK-NEXT: CXXRecordDecl{{.*}} struct S definition + // CHECK: CXXRecordDecl{{.*}} implicit struct S + int ThisMember; + // CHECK-NEXT: FieldDecl{{.*}} ThisMember 'int' + int ThisMemberArray[5]; + // CHECK-NEXT: FieldDecl{{.*}} ThisMemberArray 'int[5]' + + void foo(); + // CHECK-NEXT: CXXMethodDecl{{.*}} foo 'void ()' + + template<typename T> + void bar() { + // CHECK-NEXT: FunctionTemplateDecl{{.*}}bar + // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 0 T + // CHECK-NEXT: CXXMethodDecl{{.*}} bar 'void ()' implicit-inline + // CHECK-NEXT: CompoundStmt + +#pragma acc parallel private(ThisMember, this->ThisMemberArray[1]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this + // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay> + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(ThisMemberArray[1:2]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr{{.*}} + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(this) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + + // Check Instantiations: + // CHECK-NEXT: CXXMethodDecl{{.*}} used bar 'void ()' implicit_instantiation implicit-inline + // CHECK-NEXT: TemplateArgument type 'int' + // CHECK-NEXT: BuiltinType{{.*}} 'int' + // CHECK-NEXT: CompoundStmt + +// #pragma acc parallel private(ThisMember, this->ThisMemberArray[1]) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this + // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay> + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +// #pragma acc parallel private(ThisMemberArray[1:2]) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr{{.*}} + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +// #pragma acc parallel private(this) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt +} +}; + +void S::foo() { + // CHECK: CXXMethodDecl{{.*}} foo 'void ()' + // CHECK-NEXT: CompoundStmt +#pragma acc parallel private(ThisMember, this->ThisMemberArray[1]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this + // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay> + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(ThisMemberArray[1:2]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr{{.*}} + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(this) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: CXXThisExpr{{.*}} 'S *' this + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt +} + +template<typename U> +struct STempl { + // CHECK-NEXT: ClassTemplateDecl{{.*}} STempl + // CHECK-NEXT: TemplateTypeParmDecl{{.*}} typename depth 0 index 0 U + // CHECK-NEXT: CXXRecordDecl{{.*}} struct STempl definition + // CHECK: CXXRecordDecl{{.*}} implicit struct STempl + U ThisMember; + // CHECK-NEXT: FieldDecl{{.*}} ThisMember 'U' + U ThisMemberArray[5]; + // CHECK-NEXT: FieldDecl{{.*}} ThisMemberArray 'U[5]' + + void foo() { + // CHECK-NEXT: CXXMethodDecl {{.*}} foo 'void ()' + // CHECK-NEXT: CompoundStmt + +#pragma acc parallel private(ThisMember, this->ThisMemberArray[1]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: MemberExpr{{.*}} 'U' lvalue ->ThisMember + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this + // CHECK-NEXT: ArraySubscriptExpr{{.*}} '<dependent type>' lvalue + // CHECK-NEXT: CXXDependentScopeMemberExpr{{.*}} '<dependent type>' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(ThisMemberArray[1:2]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr{{.*}} + // CHECK-NEXT: MemberExpr{{.*}} 'U[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(this) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt +} + + template<typename T> + void bar() { + // CHECK-NEXT: FunctionTemplateDecl{{.*}} bar + // CHECK-NEXT: TemplateTypeParmDecl{{.*}} typename depth 1 index 0 T + // CHECK-NEXT: CXXMethodDecl{{.*}} bar 'void ()' + // CHECK-NEXT: CompoundStmt + +#pragma acc parallel private(ThisMember, this->ThisMemberArray[1]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: MemberExpr{{.*}} 'U' lvalue ->ThisMember + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this + // CHECK-NEXT: ArraySubscriptExpr{{.*}} '<dependent type>' lvalue + // CHECK-NEXT: CXXDependentScopeMemberExpr{{.*}} '<dependent type>' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(ThisMemberArray[1:2]) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr{{.*}} + // CHECK-NEXT: MemberExpr{{.*}} 'U[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' implicit this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +#pragma acc parallel private(this) + while(true); + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<U> *' this + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt +} + +// Instantiation of the class template. + +// CHECK-NEXT: ClassTemplateSpecializationDecl{{.*}}struct STempl +// CHECK: TemplateArgument type 'int' +// CHECK-NEXT: BuiltinType {{.*}}'int' +// CHECK-NEXT: CXXRecordDecl{{.*}} struct STempl +// CHECK-NEXT: FieldDecl{{.*}}ThisMember 'int' +// CHECK-NEXT: FieldDecl{{.*}} ThisMemberArray 'int[5]' + +// CHECK-NEXT: CXXMethodDecl{{.*}} foo 'void ()' +// CHECK-NEXT: FunctionTemplateDecl{{.*}}bar +// CHECK-NEXT: TemplateTypeParmDecl{{.*}} typename depth 0 index 0 T +// CHECK-NEXT: CXXMethodDecl{{.*}}bar 'void ()' +// CHECK-NEXT: CXXMethodDecl{{.*}}bar 'void ()' +// CHECK-NEXT: TemplateArgument type 'int' +// CHECK-NEXT: BuiltinType{{.*}} 'int' +// CHECK-NEXT: CompoundStmt + +//#pragma acc parallel private(ThisMember, this->ThisMemberArray[1]) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: MemberExpr{{.*}} 'int' lvalue ->ThisMember + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' implicit this + // CHECK-NEXT: ArraySubscriptExpr{{.*}} 'int' lvalue + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay> + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +//#pragma acc parallel private(ThisMemberArray[1:2]) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: ArraySectionExpr{{.*}} + // CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' implicit this + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 1 + // CHECK-NEXT: IntegerLiteral{{.*}}'int' 2 + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt + +//#pragma acc parallel private(this) + // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel + // CHECK-NEXT: private clause + // CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' this + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: NullStmt +}; + +void Inst() { + static constexpr unsigned CEVar = 1; + int i; + int Arr[5]; + TemplUses<CEVar, int, int[1]>({}, {}, &i); + + S s; + s.bar<int>(); + STempl<int> stempl; + stempl.bar<int>(); +} diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp index 398a11a5703558..eb0ba09c5b9116 100644 --- a/clang/tools/libclang/CIndex.cpp +++ b/clang/tools/libclang/CIndex.cpp @@ -2782,6 +2782,11 @@ class OpenACCClauseEnqueue : public OpenACCClauseVisitor<OpenACCClauseEnqueue> { public: OpenACCClauseEnqueue(EnqueueVisitor &V) : Visitor(V) {} + void VisitVarList(const OpenACCClauseWithVarList &C) { + for (Expr *Var : C.getVarList()) + Visitor.AddStmt(Var); + } + #define VISIT_CLAUSE(CLAUSE_NAME) \ void Visit##CLAUSE_NAME##Clause(const OpenACC##CLAUSE_NAME##Clause &C); #include "clang/Basic/OpenACCClauses.def" @@ -2807,6 +2812,10 @@ void OpenACCClauseEnqueue::VisitNumGangsClause(const OpenACCNumGangsClause &C) { for (Expr *IE : C.getIntExprs()) Visitor.AddStmt(IE); } + +void OpenACCClauseEnqueue::VisitPrivateClause(const OpenACCPrivateClause &C) { + VisitVarList(C); +} } // namespace void EnqueueVisitor::EnqueueChildren(const OpenACCClause *C) { _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
