[clang] b8adf16 - [OpenACC] Implement 'vector_length' clause On compute constructs
via cfe-commits
cfe-commits at lists.llvm.org
Thu Apr 18 13:27:49 PDT 2024
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 at +1{{OpenACC clause 'vector_length' not yet implemented, clause ignored}}
#pragma acc parallel vector_length(5)
{}
- // expected-warning at +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 at +1{{OpenACC clause 'vector_length' not yet implemented, clause ignored}}
#pragma acc parallel vector_length(T::value)
for(;;){}
- // expected-warning at +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 at +1{{OpenACC 'vector_length' clause is not valid on 'serial' directive}}
+#pragma acc serial vector_length(1)
+ while(1);
+
+ struct NotConvertible{} NC;
+ // expected-error at +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 at +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 at +1{{OpenACC clause 'vector_length' requires expression of integer type ('struct NotConvertible' invalid}}
+#pragma acc parallel vector_length(NC)
+ while(1);
+
+ // expected-error at +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 at +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 at +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 at +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 at +1{{no member named 'Invalid' in 'HasInt'}}
+#pragma acc parallel vector_length(HasInt::Invalid)
+ while (1);
+
+ // expected-error at +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 at +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 at +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 at +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 at +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) {
More information about the cfe-commits
mailing list