Author: Erich Keane Date: 2024-04-18T13:27:42-07:00 New Revision: b8adf169bb86f8226978e1262443e2ffd94298ce
URL: https://github.com/llvm/llvm-project/commit/b8adf169bb86f8226978e1262443e2ffd94298ce DIFF: https://github.com/llvm/llvm-project/commit/b8adf169bb86f8226978e1262443e2ffd94298ce.diff LOG: [OpenACC] Implement 'vector_length' clause On compute constructs The 'vector_length' clause is semantically identical to the 'num_workers' clause, in that it takes a mandatory single int-expr. This is implemented identically to it. Added: clang/test/SemaOpenACC/compute-construct-vector_length-clause.c clang/test/SemaOpenACC/compute-construct-vector_length-clause.cpp Modified: clang/include/clang/AST/OpenACCClause.h clang/include/clang/Basic/OpenACCClauses.def clang/include/clang/Sema/SemaOpenACC.h clang/lib/AST/OpenACCClause.cpp clang/lib/AST/StmtProfile.cpp clang/lib/AST/TextNodeDumper.cpp clang/lib/Parse/ParseOpenACC.cpp clang/lib/Sema/SemaOpenACC.cpp clang/lib/Sema/TreeTransform.h clang/lib/Serialization/ASTReader.cpp clang/lib/Serialization/ASTWriter.cpp clang/test/ParserOpenACC/parse-clauses.c clang/test/ParserOpenACC/parse-clauses.cpp clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp clang/tools/libclang/CIndex.cpp Removed: ################################################################################ diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h index a4cde9ae8a6650..8b6d3221aa066b 100644 --- a/clang/include/clang/AST/OpenACCClause.h +++ b/clang/include/clang/AST/OpenACCClause.h @@ -196,6 +196,16 @@ class OpenACCNumWorkersClause : public OpenACCClauseWithSingleIntExpr { Expr *IntExpr, SourceLocation EndLoc); }; +class OpenACCVectorLengthClause : public OpenACCClauseWithSingleIntExpr { + OpenACCVectorLengthClause(SourceLocation BeginLoc, SourceLocation LParenLoc, + Expr *IntExpr, SourceLocation EndLoc); + +public: + static OpenACCVectorLengthClause * + Create(const ASTContext &C, SourceLocation BeginLoc, SourceLocation LParenLoc, + Expr *IntExpr, SourceLocation EndLoc); +}; + template <class Impl> class OpenACCClauseVisitor { Impl &getDerived() { return static_cast<Impl &>(*this); } diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def index d1a95cbe613944..520e068f4ffd40 100644 --- a/clang/include/clang/Basic/OpenACCClauses.def +++ b/clang/include/clang/Basic/OpenACCClauses.def @@ -19,5 +19,6 @@ VISIT_CLAUSE(Default) VISIT_CLAUSE(If) VISIT_CLAUSE(Self) VISIT_CLAUSE(NumWorkers) +VISIT_CLAUSE(VectorLength) #undef VISIT_CLAUSE diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h index b995edebe8eed3..023722049732af 100644 --- a/clang/include/clang/Sema/SemaOpenACC.h +++ b/clang/include/clang/Sema/SemaOpenACC.h @@ -93,13 +93,15 @@ class SemaOpenACC : public SemaBase { } unsigned getNumIntExprs() const { - assert(ClauseKind == OpenACCClauseKind::NumWorkers && + assert((ClauseKind == OpenACCClauseKind::NumWorkers || + ClauseKind == OpenACCClauseKind::VectorLength) && "Parsed clause kind does not have a int exprs"); return std::get<IntExprDetails>(Details).IntExprs.size(); } ArrayRef<Expr *> getIntExprs() { - assert(ClauseKind == OpenACCClauseKind::NumWorkers && + assert((ClauseKind == OpenACCClauseKind::NumWorkers || + ClauseKind == OpenACCClauseKind::VectorLength) && "Parsed clause kind does not have a int exprs"); return std::get<IntExprDetails>(Details).IntExprs; } @@ -132,7 +134,8 @@ class SemaOpenACC : public SemaBase { } void setIntExprDetails(ArrayRef<Expr *> IntExprs) { - assert(ClauseKind == OpenACCClauseKind::NumWorkers && + assert((ClauseKind == OpenACCClauseKind::NumWorkers || + ClauseKind == OpenACCClauseKind::VectorLength) && "Parsed clause kind does not have a int exprs"); Details = IntExprDetails{{IntExprs.begin(), IntExprs.end()}}; } diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp index 3fff02fa33f28d..75334223e073c3 100644 --- a/clang/lib/AST/OpenACCClause.cpp +++ b/clang/lib/AST/OpenACCClause.cpp @@ -103,6 +103,27 @@ OpenACCNumWorkersClause::Create(const ASTContext &C, SourceLocation BeginLoc, OpenACCNumWorkersClause(BeginLoc, LParenLoc, IntExpr, EndLoc); } +OpenACCVectorLengthClause::OpenACCVectorLengthClause(SourceLocation BeginLoc, + SourceLocation LParenLoc, + Expr *IntExpr, + SourceLocation EndLoc) + : OpenACCClauseWithSingleIntExpr(OpenACCClauseKind::VectorLength, BeginLoc, + LParenLoc, IntExpr, EndLoc) { + assert((!IntExpr || IntExpr->isInstantiationDependent() || + IntExpr->getType()->isIntegerType()) && + "Condition expression type not scalar/dependent"); +} + +OpenACCVectorLengthClause * +OpenACCVectorLengthClause::Create(const ASTContext &C, SourceLocation BeginLoc, + SourceLocation LParenLoc, Expr *IntExpr, + SourceLocation EndLoc) { + void *Mem = C.Allocate(sizeof(OpenACCVectorLengthClause), + alignof(OpenACCVectorLengthClause)); + return new (Mem) + OpenACCVectorLengthClause(BeginLoc, LParenLoc, IntExpr, EndLoc); +} + //===----------------------------------------------------------------------===// // OpenACC clauses printing methods //===----------------------------------------------------------------------===// @@ -124,3 +145,8 @@ void OpenACCClausePrinter::VisitNumWorkersClause( const OpenACCNumWorkersClause &C) { OS << "num_workers(" << C.getIntExpr() << ")"; } + +void OpenACCClausePrinter::VisitVectorLengthClause( + const OpenACCVectorLengthClause &C) { + OS << "vector_length(" << C.getIntExpr() << ")"; +} diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp index cc230909bd3840..8138ae3a244a8b 100644 --- a/clang/lib/AST/StmtProfile.cpp +++ b/clang/lib/AST/StmtProfile.cpp @@ -2503,6 +2503,12 @@ void OpenACCClauseProfiler::VisitNumWorkersClause( Profiler.VisitStmt(Clause.getIntExpr()); } +void OpenACCClauseProfiler::VisitVectorLengthClause( + const OpenACCVectorLengthClause &Clause) { + assert(Clause.hasIntExpr() && + "vector_length clause requires a valid int expr"); + Profiler.VisitStmt(Clause.getIntExpr()); +} } // namespace void StmtProfiler::VisitOpenACCComputeConstruct( diff --git a/clang/lib/AST/TextNodeDumper.cpp b/clang/lib/AST/TextNodeDumper.cpp index 9d1b73cb7a0784..e5a8b285715b30 100644 --- a/clang/lib/AST/TextNodeDumper.cpp +++ b/clang/lib/AST/TextNodeDumper.cpp @@ -400,6 +400,7 @@ void TextNodeDumper::Visit(const OpenACCClause *C) { case OpenACCClauseKind::If: case OpenACCClauseKind::Self: case OpenACCClauseKind::NumWorkers: + 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. OS << " clause"; diff --git a/clang/lib/Parse/ParseOpenACC.cpp b/clang/lib/Parse/ParseOpenACC.cpp index 096e0863ed47c7..757417f75c9636 100644 --- a/clang/lib/Parse/ParseOpenACC.cpp +++ b/clang/lib/Parse/ParseOpenACC.cpp @@ -960,7 +960,8 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams( // TODO OpenACC: as we implement the 'rest' of the above, this 'if' should // be removed leaving just the 'setIntExprDetails'. - if (ClauseKind == OpenACCClauseKind::NumWorkers) + if (ClauseKind == OpenACCClauseKind::NumWorkers || + ClauseKind == OpenACCClauseKind::VectorLength) ParsedClause.setIntExprDetails(IntExpr.get()); break; diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp index 39cde677ecc87a..190739fa02e932 100644 --- a/clang/lib/Sema/SemaOpenACC.cpp +++ b/clang/lib/Sema/SemaOpenACC.cpp @@ -92,6 +92,7 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind, return false; } case OpenACCClauseKind::NumWorkers: + case OpenACCClauseKind::VectorLength: switch (DirectiveKind) { case OpenACCDirectiveKind::Parallel: case OpenACCDirectiveKind::Kernels: @@ -248,6 +249,25 @@ SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses, getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(), Clause.getIntExprs()[0], Clause.getEndLoc()); } + case OpenACCClauseKind::VectorLength: { + // 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; + + // 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. + if (checkAlreadyHasClauseOfKind(*this, ExistingClauses, Clause)) + return nullptr; + + assert(Clause.getIntExprs().size() == 1 && + "Invalid number of expressions for VectorLength"); + return OpenACCVectorLengthClause::Create( + getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(), + Clause.getIntExprs()[0], Clause.getEndLoc()); + } default: break; } diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h index 34b52a346a10d1..ade33ec65038fd 100644 --- a/clang/lib/Sema/TreeTransform.h +++ b/clang/lib/Sema/TreeTransform.h @@ -11181,6 +11181,29 @@ void OpenACCClauseTransform<Derived>::VisitNumWorkersClause( ParsedClause.getLParenLoc(), ParsedClause.getIntExprs()[0], ParsedClause.getEndLoc()); } + +template <typename Derived> +void OpenACCClauseTransform<Derived>::VisitVectorLengthClause( + const OpenACCVectorLengthClause &C) { + Expr *IntExpr = const_cast<Expr *>(C.getIntExpr()); + assert(IntExpr && "vector_length clause constructed with invalid int expr"); + + ExprResult Res = Self.TransformExpr(IntExpr); + if (!Res.isUsable()) + return; + + Res = Self.getSema().OpenACC().ActOnIntExpr(OpenACCDirectiveKind::Invalid, + C.getClauseKind(), + C.getBeginLoc(), Res.get()); + if (!Res.isUsable()) + return; + + ParsedClause.setIntExprDetails(Res.get()); + NewClause = OpenACCVectorLengthClause::Create( + Self.getSema().getASTContext(), ParsedClause.getBeginLoc(), + ParsedClause.getLParenLoc(), ParsedClause.getIntExprs()[0], + ParsedClause.getEndLoc()); +} } // namespace template <typename Derived> OpenACCClause *TreeTransform<Derived>::TransformOpenACCClause( diff --git a/clang/lib/Serialization/ASTReader.cpp b/clang/lib/Serialization/ASTReader.cpp index 6fff829cefb845..44e23919ea18e0 100644 --- a/clang/lib/Serialization/ASTReader.cpp +++ b/clang/lib/Serialization/ASTReader.cpp @@ -11792,6 +11792,12 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { return OpenACCNumWorkersClause::Create(getContext(), BeginLoc, LParenLoc, IntExpr, EndLoc); } + case OpenACCClauseKind::VectorLength: { + SourceLocation LParenLoc = readSourceLocation(); + Expr *IntExpr = readSubExpr(); + return OpenACCVectorLengthClause::Create(getContext(), BeginLoc, LParenLoc, + IntExpr, EndLoc); + } case OpenACCClauseKind::Finalize: case OpenACCClauseKind::IfPresent: case OpenACCClauseKind::Seq: @@ -11820,7 +11826,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() { case OpenACCClauseKind::Reduction: case OpenACCClauseKind::Collapse: case OpenACCClauseKind::Bind: - case OpenACCClauseKind::VectorLength: case OpenACCClauseKind::NumGangs: case OpenACCClauseKind::DeviceNum: case OpenACCClauseKind::DefaultAsync: diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp index 934b202ac38267..6dd87b5d200db6 100644 --- a/clang/lib/Serialization/ASTWriter.cpp +++ b/clang/lib/Serialization/ASTWriter.cpp @@ -5014,7 +5014,7 @@ void ASTWriter::WriteSpecialDeclRecords(Sema &SemaRef) { if (!ModularCodegenDecls.empty()) Stream.EmitRecord(MODULAR_CODEGEN_DECLS, ModularCodegenDecls); - + // Write the record containing tentative definitions. RecordData TentativeDefinitions; AddLazyVectorEmiitedDecls(*this, SemaRef.TentativeDefinitions, @@ -5135,7 +5135,7 @@ void ASTWriter::WriteSpecialDeclRecords(Sema &SemaRef) { } if (!UndefinedButUsed.empty()) Stream.EmitRecord(UNDEFINED_BUT_USED, UndefinedButUsed); - + // Write all delete-expressions that we would like to // analyze later in AST. RecordData DeleteExprsToAnalyze; @@ -7663,6 +7663,12 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { AddStmt(const_cast<Expr *>(NWC->getIntExpr())); return; } + case OpenACCClauseKind::VectorLength: { + const auto *NWC = cast<OpenACCVectorLengthClause>(C); + writeSourceLocation(NWC->getLParenLoc()); + AddStmt(const_cast<Expr *>(NWC->getIntExpr())); + return; + } case OpenACCClauseKind::Finalize: case OpenACCClauseKind::IfPresent: case OpenACCClauseKind::Seq: @@ -7691,7 +7697,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) { case OpenACCClauseKind::Reduction: case OpenACCClauseKind::Collapse: case OpenACCClauseKind::Bind: - case OpenACCClauseKind::VectorLength: case OpenACCClauseKind::NumGangs: case OpenACCClauseKind::DeviceNum: case OpenACCClauseKind::DefaultAsync: diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c index bd5f459eae074f..ddf40f71701eda 100644 --- a/clang/test/ParserOpenACC/parse-clauses.c +++ b/clang/test/ParserOpenACC/parse-clauses.c @@ -893,11 +893,9 @@ void IntExprParsing() { #pragma acc parallel vector_length(5, 4) {} - // expected-warning@+1{{OpenACC clause 'vector_length' not yet implemented, clause ignored}} #pragma acc parallel vector_length(5) {} - // expected-warning@+1{{OpenACC clause 'vector_length' not yet implemented, clause ignored}} #pragma acc parallel vector_length(returns_int()) {} diff --git a/clang/test/ParserOpenACC/parse-clauses.cpp b/clang/test/ParserOpenACC/parse-clauses.cpp index 09a90726c69396..8c1d6437479961 100644 --- a/clang/test/ParserOpenACC/parse-clauses.cpp +++ b/clang/test/ParserOpenACC/parse-clauses.cpp @@ -12,11 +12,9 @@ void templ() { #pragma acc loop collapse(T::value) for(;;){} - // expected-warning@+1{{OpenACC clause 'vector_length' not yet implemented, clause ignored}} #pragma acc parallel vector_length(T::value) for(;;){} - // expected-warning@+1{{OpenACC clause 'vector_length' not yet implemented, clause ignored}} #pragma acc parallel vector_length(I) for(;;){} diff --git a/clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp b/clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp index e3664460a6e5e3..889025c26818d8 100644 --- a/clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp +++ b/clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp @@ -77,6 +77,17 @@ void NormalUses() { // CHECK-NEXT: WhileStmt // CHECK-NEXT: CXXBoolLiteralExpr // CHECK-NEXT: CompoundStmt + +#pragma acc kernels vector_length(some_short()) + while(true){} + // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels + // CHECK-NEXT: vector_length clause + // CHECK-NEXT: CallExpr{{.*}}'short' + // CHECK-NEXT: ImplicitCastExpr{{.*}}'short (*)()' <FunctionToPointerDecay> + // CHECK-NEXT: DeclRefExpr{{.*}}'short ()' lvalue Function{{.*}} 'some_short' 'short ()' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: CompoundStmt } template<typename T, typename U> @@ -157,6 +168,25 @@ void TemplUses(T t, U u) { // CHECK-NEXT: CXXBoolLiteralExpr // CHECK-NEXT: CompoundStmt +#pragma acc kernels vector_length(u) + while(true){} + // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels + // CHECK-NEXT: vector_length clause + // CHECK-NEXT: DeclRefExpr{{.*}} 'U' lvalue ParmVar{{.*}} 'u' 'U' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: CompoundStmt + +#pragma acc parallel vector_length(U::value) + while(true){} + // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel + // CHECK-NEXT: vector_length clause + // CHECK-NEXT: DependentScopeDeclRefExpr{{.*}} '<dependent type>' lvalue + // CHECK-NEXT: NestedNameSpecifier TypeSpec 'U' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: CompoundStmt + // Check the instantiated versions of the above. // CHECK-NEXT: FunctionDecl{{.*}} used TemplUses 'void (CorrectConvert, HasInt)' implicit_instantiation // CHECK-NEXT: TemplateArgument type 'CorrectConvert' @@ -239,6 +269,25 @@ void TemplUses(T t, U u) { // CHECK-NEXT: WhileStmt // CHECK-NEXT: CXXBoolLiteralExpr // CHECK-NEXT: CompoundStmt + + // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels + // CHECK-NEXT: vector_length clause + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'char' <UserDefinedConversion> + // CHECK-NEXT: CXXMemberCallExpr{{.*}}'char' + // CHECK-NEXT: MemberExpr{{.*}} '<bound member function type>' .operator char + // CHECK-NEXT: DeclRefExpr{{.*}} 'HasInt' lvalue ParmVar + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: CompoundStmt + + // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel + // CHECK-NEXT: vector_length clause + // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue> + // CHECK-NEXT: DeclRefExpr{{.*}} 'const int' lvalue Var{{.*}} 'value' 'const int' + // CHECK-NEXT: NestedNameSpecifier TypeSpec 'HasInt' + // CHECK-NEXT: WhileStmt + // CHECK-NEXT: CXXBoolLiteralExpr + // CHECK-NEXT: CompoundStmt } struct HasInt { diff --git a/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c new file mode 100644 index 00000000000000..cd85bdefb602d3 --- /dev/null +++ b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c @@ -0,0 +1,33 @@ +// RUN: %clang_cc1 %s -fopenacc -verify + +short getS(); + +void Test() { +#pragma acc parallel vector_length(1) + while(1); +#pragma acc kernels vector_length(1) + while(1); + + // expected-error@+1{{OpenACC 'vector_length' clause is not valid on 'serial' directive}} +#pragma acc serial vector_length(1) + while(1); + + struct NotConvertible{} NC; + // expected-error@+1{{OpenACC clause 'vector_length' requires expression of integer type ('struct NotConvertible' invalid)}} +#pragma acc parallel vector_length(NC) + while(1); + +#pragma acc kernels vector_length(getS()) + while(1); + + struct Incomplete *SomeIncomplete; + + // expected-error@+1{{OpenACC clause 'vector_length' requires expression of integer type ('struct Incomplete' invalid)}} +#pragma acc kernels vector_length(*SomeIncomplete) + while(1); + + enum E{A} SomeE; + +#pragma acc kernels vector_length(SomeE) + while(1); +} diff --git a/clang/test/SemaOpenACC/compute-construct-vector_length-clause.cpp b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.cpp new file mode 100644 index 00000000000000..f6c5dde1a02355 --- /dev/null +++ b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.cpp @@ -0,0 +1,133 @@ +// RUN: %clang_cc1 %s -fopenacc -verify + +struct NotConvertible{} NC; +struct Incomplete *SomeIncomplete; // #INCOMPLETE +enum E{} SomeE; +enum class E2{} SomeE2; + +struct CorrectConvert { + operator int(); +} Convert; + +struct ExplicitConvertOnly { + explicit operator int() const; // #EXPL_CONV +} Explicit; + +struct AmbiguousConvert{ + operator int(); // #AMBIG_INT + operator short(); // #AMBIG_SHORT + operator float(); +} Ambiguous; + +void Test() { +#pragma acc parallel vector_length(1) + while(1); +#pragma acc kernels vector_length(1) + while(1); + + // expected-error@+1{{OpenACC clause 'vector_length' requires expression of integer type ('struct NotConvertible' invalid}} +#pragma acc parallel vector_length(NC) + while(1); + + // expected-error@+2{{OpenACC integer expression has incomplete class type 'struct Incomplete'}} + // expected-note@#INCOMPLETE{{forward declaration of 'Incomplete'}} +#pragma acc kernels vector_length(*SomeIncomplete) + while(1); + +#pragma acc parallel vector_length(SomeE) + while(1); + + // expected-error@+1{{OpenACC clause 'vector_length' requires expression of integer type ('enum E2' invalid}} +#pragma acc kernels vector_length(SomeE2) + while(1); + +#pragma acc parallel vector_length(Convert) + while(1); + + // expected-error@+2{{OpenACC integer expression type 'struct ExplicitConvertOnly' requires explicit conversion to 'int'}} + // expected-note@#EXPL_CONV{{conversion to integral type 'int'}} +#pragma acc kernels vector_length(Explicit) + while(1); + + // expected-error@+3{{multiple conversions from expression type 'struct AmbiguousConvert' to an integral type}} + // expected-note@#AMBIG_INT{{conversion to integral type 'int'}} + // expected-note@#AMBIG_SHORT{{conversion to integral type 'short'}} +#pragma acc parallel vector_length(Ambiguous) + while(1); +} + +struct HasInt { + using IntTy = int; + using ShortTy = short; + static constexpr int value = 1; + static constexpr AmbiguousConvert ACValue; + static constexpr ExplicitConvertOnly EXValue; + + operator char(); +}; + +template<typename T> +void TestInst() { + + // expected-error@+1{{no member named 'Invalid' in 'HasInt'}} +#pragma acc parallel vector_length(HasInt::Invalid) + while (1); + + // expected-error@+2{{no member named 'Invalid' in 'HasInt'}} + // expected-note@#INST{{in instantiation of function template specialization 'TestInst<HasInt>' requested here}} +#pragma acc kernels vector_length(T::Invalid) + while (1); + + // expected-error@+3{{multiple conversions from expression type 'const AmbiguousConvert' to an integral type}} + // expected-note@#AMBIG_INT{{conversion to integral type 'int'}} + // expected-note@#AMBIG_SHORT{{conversion to integral type 'short'}} +#pragma acc parallel vector_length(HasInt::ACValue) + while (1); + + // expected-error@+3{{multiple conversions from expression type 'const AmbiguousConvert' to an integral type}} + // expected-note@#AMBIG_INT{{conversion to integral type 'int'}} + // expected-note@#AMBIG_SHORT{{conversion to integral type 'short'}} +#pragma acc kernels vector_length(T::ACValue) + while (1); + + // expected-error@+2{{OpenACC integer expression type 'const ExplicitConvertOnly' requires explicit conversion to 'int'}} + // expected-note@#EXPL_CONV{{conversion to integral type 'int'}} +#pragma acc parallel vector_length(HasInt::EXValue) + while (1); + + // expected-error@+2{{OpenACC integer expression type 'const ExplicitConvertOnly' requires explicit conversion to 'int'}} + // expected-note@#EXPL_CONV{{conversion to integral type 'int'}} +#pragma acc kernels vector_length(T::EXValue) + while (1); + +#pragma acc parallel vector_length(HasInt::value) + while (1); + +#pragma acc kernels vector_length(T::value) + while (1); + +#pragma acc parallel vector_length(HasInt::IntTy{}) + while (1); + +#pragma acc kernels vector_length(typename T::ShortTy{}) + while (1); + +#pragma acc parallel vector_length(HasInt::IntTy{}) + while (1); + +#pragma acc kernels vector_length(typename T::ShortTy{}) + while (1); + + HasInt HI{}; + T MyT{}; + +#pragma acc parallel vector_length(HI) + while (1); + +#pragma acc kernels vector_length(MyT) + while (1); +} + +void Inst() { + TestInst<HasInt>(); // #INST +} diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp index 011bec32bab319..cbc1d85bb33dfc 100644 --- a/clang/tools/libclang/CIndex.cpp +++ b/clang/tools/libclang/CIndex.cpp @@ -2799,6 +2799,10 @@ void OpenACCClauseEnqueue::VisitNumWorkersClause( const OpenACCNumWorkersClause &C) { Visitor.AddStmt(C.getIntExpr()); } +void OpenACCClauseEnqueue::VisitVectorLengthClause( + const OpenACCVectorLengthClause &C) { + Visitor.AddStmt(C.getIntExpr()); +} } // 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