[clang] 76600ae - [OpenACC] Implement 'num_workers' clause for compute constructs (#89151)
via cfe-commits
cfe-commits at lists.llvm.org
Thu Apr 18 12:42:26 PDT 2024
Author: Erich Keane
Date: 2024-04-18T12:42:22-07:00
New Revision: 76600aee9d7c851d3e26eed7496cec880bd53861
URL: https://github.com/llvm/llvm-project/commit/76600aee9d7c851d3e26eed7496cec880bd53861
DIFF: https://github.com/llvm/llvm-project/commit/76600aee9d7c851d3e26eed7496cec880bd53861.diff
LOG: [OpenACC] Implement 'num_workers' clause for compute constructs (#89151)
This clause just takes an 'int expr', which is not optional. This patch
implements the clause on compute constructs.
Added:
clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp
clang/test/SemaOpenACC/compute-construct-num_workers-clause.c
clang/test/SemaOpenACC/compute-construct-num_workers-clause.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/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/tools/libclang/CIndex.cpp
Removed:
################################################################################
diff --git a/clang/include/clang/AST/OpenACCClause.h b/clang/include/clang/AST/OpenACCClause.h
index 07587849eb1219..a4cde9ae8a6650 100644
--- a/clang/include/clang/AST/OpenACCClause.h
+++ b/clang/include/clang/AST/OpenACCClause.h
@@ -156,6 +156,46 @@ class OpenACCSelfClause : public OpenACCClauseWithCondition {
Expr *ConditionExpr, SourceLocation EndLoc);
};
+/// Represents oen of a handful of classes that have a single integer
+/// expression.
+class OpenACCClauseWithSingleIntExpr : public OpenACCClauseWithParams {
+ Expr *IntExpr;
+
+protected:
+ OpenACCClauseWithSingleIntExpr(OpenACCClauseKind K, SourceLocation BeginLoc,
+ SourceLocation LParenLoc, Expr *IntExpr,
+ SourceLocation EndLoc)
+ : OpenACCClauseWithParams(K, BeginLoc, LParenLoc, EndLoc),
+ IntExpr(IntExpr) {}
+
+public:
+ bool hasIntExpr() const { return IntExpr; }
+ const Expr *getIntExpr() const { return IntExpr; }
+
+ Expr *getIntExpr() { return IntExpr; };
+
+ child_range children() {
+ return child_range(reinterpret_cast<Stmt **>(&IntExpr),
+ reinterpret_cast<Stmt **>(&IntExpr + 1));
+ }
+
+ const_child_range children() const {
+ return const_child_range(reinterpret_cast<Stmt *const *>(&IntExpr),
+ reinterpret_cast<Stmt *const *>(&IntExpr + 1));
+ }
+};
+
+class OpenACCNumWorkersClause : public OpenACCClauseWithSingleIntExpr {
+ OpenACCNumWorkersClause(SourceLocation BeginLoc, SourceLocation LParenLoc,
+ Expr *IntExpr, SourceLocation EndLoc);
+
+public:
+ static OpenACCNumWorkersClause *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/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index a9f4143c6b375e..1a2d8bf4e4eb15 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12271,4 +12271,16 @@ def warn_acc_if_self_conflict
: Warning<"OpenACC construct 'self' has no effect when an 'if' clause "
"evaluates to true">,
InGroup<DiagGroup<"openacc-self-if-potential-conflict">>;
+def err_acc_int_expr_requires_integer
+ : Error<"OpenACC %select{clause|directive}0 '%1' requires expression of "
+ "integer type (%2 invalid)">;
+def err_acc_int_expr_incomplete_class_type
+ : Error<"OpenACC integer expression has incomplete class type %0">;
+def err_acc_int_expr_explicit_conversion
+ : Error<"OpenACC integer expression type %0 requires explicit conversion "
+ "to %1">;
+def note_acc_int_expr_conversion
+ : Note<"conversion to %select{integral|enumeration}0 type %1">;
+def err_acc_int_expr_multiple_conversions
+ : Error<"multiple conversions from expression type %0 to an integral type">;
} // end of sema component.
diff --git a/clang/include/clang/Basic/OpenACCClauses.def b/clang/include/clang/Basic/OpenACCClauses.def
index 378495d2c0909a..d1a95cbe613944 100644
--- a/clang/include/clang/Basic/OpenACCClauses.def
+++ b/clang/include/clang/Basic/OpenACCClauses.def
@@ -18,5 +18,6 @@
VISIT_CLAUSE(Default)
VISIT_CLAUSE(If)
VISIT_CLAUSE(Self)
+VISIT_CLAUSE(NumWorkers)
#undef VISIT_CLAUSE
diff --git a/clang/include/clang/Parse/Parser.h b/clang/include/clang/Parse/Parser.h
index 23b268126de4e0..72b2f958a5e622 100644
--- a/clang/include/clang/Parse/Parser.h
+++ b/clang/include/clang/Parse/Parser.h
@@ -3640,13 +3640,14 @@ class Parser : public CodeCompletionHandler {
/// Parses the clause-list for an OpenACC directive.
SmallVector<OpenACCClause *>
ParseOpenACCClauseList(OpenACCDirectiveKind DirKind);
- bool ParseOpenACCWaitArgument();
+ bool ParseOpenACCWaitArgument(SourceLocation Loc, bool IsDirective);
/// Parses the clause of the 'bind' argument, which can be a string literal or
/// an ID expression.
ExprResult ParseOpenACCBindClauseArgument();
/// Parses the clause kind of 'int-expr', which can be any integral
/// expression.
- ExprResult ParseOpenACCIntExpr();
+ ExprResult ParseOpenACCIntExpr(OpenACCDirectiveKind DK, OpenACCClauseKind CK,
+ SourceLocation Loc);
/// Parses the 'device-type-list', which is a list of identifiers.
bool ParseOpenACCDeviceTypeList();
/// Parses the 'async-argument', which is an integral value with two
@@ -3657,9 +3658,9 @@ class Parser : public CodeCompletionHandler {
/// Parses a comma delimited list of 'size-expr's.
bool ParseOpenACCSizeExprList();
/// Parses a 'gang-arg-list', used for the 'gang' clause.
- bool ParseOpenACCGangArgList();
+ bool ParseOpenACCGangArgList(SourceLocation GangLoc);
/// Parses a 'gang-arg', used for the 'gang' clause.
- bool ParseOpenACCGangArg();
+ bool ParseOpenACCGangArg(SourceLocation GangLoc);
/// Parses a 'condition' expr, ensuring it results in a
ExprResult ParseOpenACCConditionExpr();
diff --git a/clang/include/clang/Sema/SemaOpenACC.h b/clang/include/clang/Sema/SemaOpenACC.h
index 329dc3945fa2a6..b995edebe8eed3 100644
--- a/clang/include/clang/Sema/SemaOpenACC.h
+++ b/clang/include/clang/Sema/SemaOpenACC.h
@@ -44,8 +44,13 @@ class SemaOpenACC : public SemaBase {
Expr *ConditionExpr;
};
- std::variant<std::monostate, DefaultDetails, ConditionDetails> Details =
- std::monostate{};
+ struct IntExprDetails {
+ SmallVector<Expr *> IntExprs;
+ };
+
+ std::variant<std::monostate, DefaultDetails, ConditionDetails,
+ IntExprDetails>
+ Details = std::monostate{};
public:
OpenACCParsedClause(OpenACCDirectiveKind DirKind,
@@ -87,6 +92,22 @@ class SemaOpenACC : public SemaBase {
return std::get<ConditionDetails>(Details).ConditionExpr;
}
+ unsigned getNumIntExprs() const {
+ assert(ClauseKind == OpenACCClauseKind::NumWorkers &&
+ "Parsed clause kind does not have a int exprs");
+ return std::get<IntExprDetails>(Details).IntExprs.size();
+ }
+
+ ArrayRef<Expr *> getIntExprs() {
+ assert(ClauseKind == OpenACCClauseKind::NumWorkers &&
+ "Parsed clause kind does not have a int exprs");
+ return std::get<IntExprDetails>(Details).IntExprs;
+ }
+
+ ArrayRef<Expr *> getIntExprs() const {
+ return const_cast<OpenACCParsedClause *>(this)->getIntExprs();
+ }
+
void setLParenLoc(SourceLocation EndLoc) { LParenLoc = EndLoc; }
void setEndLoc(SourceLocation EndLoc) { ClauseRange.setEnd(EndLoc); }
@@ -109,6 +130,12 @@ class SemaOpenACC : public SemaBase {
Details = ConditionDetails{ConditionExpr};
}
+
+ void setIntExprDetails(ArrayRef<Expr *> IntExprs) {
+ assert(ClauseKind == OpenACCClauseKind::NumWorkers &&
+ "Parsed clause kind does not have a int exprs");
+ Details = IntExprDetails{{IntExprs.begin(), IntExprs.end()}};
+ }
};
SemaOpenACC(Sema &S);
@@ -148,6 +175,11 @@ class SemaOpenACC : public SemaBase {
/// Called after the directive has been completely parsed, including the
/// declaration group or associated statement.
DeclGroupRef ActOnEndDeclDirective();
+
+ /// Called when encountering an 'int-expr' for OpenACC, and manages
+ /// conversions and diagnostics to 'int'.
+ ExprResult ActOnIntExpr(OpenACCDirectiveKind DK, OpenACCClauseKind CK,
+ SourceLocation Loc, Expr *IntExpr);
};
} // namespace clang
diff --git a/clang/lib/AST/OpenACCClause.cpp b/clang/lib/AST/OpenACCClause.cpp
index 9c259c8f9bd0a1..3fff02fa33f28d 100644
--- a/clang/lib/AST/OpenACCClause.cpp
+++ b/clang/lib/AST/OpenACCClause.cpp
@@ -82,6 +82,27 @@ OpenACCClause::child_range OpenACCClause::children() {
return child_range(child_iterator(), child_iterator());
}
+OpenACCNumWorkersClause::OpenACCNumWorkersClause(SourceLocation BeginLoc,
+ SourceLocation LParenLoc,
+ Expr *IntExpr,
+ SourceLocation EndLoc)
+ : OpenACCClauseWithSingleIntExpr(OpenACCClauseKind::NumWorkers, BeginLoc,
+ LParenLoc, IntExpr, EndLoc) {
+ assert((!IntExpr || IntExpr->isInstantiationDependent() ||
+ IntExpr->getType()->isIntegerType()) &&
+ "Condition expression type not scalar/dependent");
+}
+
+OpenACCNumWorkersClause *
+OpenACCNumWorkersClause::Create(const ASTContext &C, SourceLocation BeginLoc,
+ SourceLocation LParenLoc, Expr *IntExpr,
+ SourceLocation EndLoc) {
+ void *Mem = C.Allocate(sizeof(OpenACCNumWorkersClause),
+ alignof(OpenACCNumWorkersClause));
+ return new (Mem)
+ OpenACCNumWorkersClause(BeginLoc, LParenLoc, IntExpr, EndLoc);
+}
+
//===----------------------------------------------------------------------===//
// OpenACC clauses printing methods
//===----------------------------------------------------------------------===//
@@ -98,3 +119,8 @@ void OpenACCClausePrinter::VisitSelfClause(const OpenACCSelfClause &C) {
if (const Expr *CondExpr = C.getConditionExpr())
OS << "(" << CondExpr << ")";
}
+
+void OpenACCClausePrinter::VisitNumWorkersClause(
+ const OpenACCNumWorkersClause &C) {
+ OS << "num_workers(" << C.getIntExpr() << ")";
+}
diff --git a/clang/lib/AST/StmtProfile.cpp b/clang/lib/AST/StmtProfile.cpp
index b26d804c6f079b..cc230909bd3840 100644
--- a/clang/lib/AST/StmtProfile.cpp
+++ b/clang/lib/AST/StmtProfile.cpp
@@ -2496,6 +2496,13 @@ void OpenACCClauseProfiler::VisitSelfClause(const OpenACCSelfClause &Clause) {
if (Clause.hasConditionExpr())
Profiler.VisitStmt(Clause.getConditionExpr());
}
+
+void OpenACCClauseProfiler::VisitNumWorkersClause(
+ const OpenACCNumWorkersClause &Clause) {
+ assert(Clause.hasIntExpr() && "num_workers 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 ff5b3df2d6dfac..9d1b73cb7a0784 100644
--- a/clang/lib/AST/TextNodeDumper.cpp
+++ b/clang/lib/AST/TextNodeDumper.cpp
@@ -399,6 +399,7 @@ void TextNodeDumper::Visit(const OpenACCClause *C) {
break;
case OpenACCClauseKind::If:
case OpenACCClauseKind::Self:
+ case OpenACCClauseKind::NumWorkers:
// 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 123be476e928ee..096e0863ed47c7 100644
--- a/clang/lib/Parse/ParseOpenACC.cpp
+++ b/clang/lib/Parse/ParseOpenACC.cpp
@@ -632,10 +632,16 @@ Parser::ParseOpenACCClauseList(OpenACCDirectiveKind DirKind) {
return Clauses;
}
-ExprResult Parser::ParseOpenACCIntExpr() {
- // FIXME: this is required to be an integer expression (or dependent), so we
- // should ensure that is the case by passing this to SEMA here.
- return getActions().CorrectDelayedTyposInExpr(ParseAssignmentExpression());
+ExprResult Parser::ParseOpenACCIntExpr(OpenACCDirectiveKind DK,
+ OpenACCClauseKind CK,
+ SourceLocation Loc) {
+ ExprResult ER =
+ getActions().CorrectDelayedTyposInExpr(ParseAssignmentExpression());
+
+ if (!ER.isUsable())
+ return ER;
+
+ return getActions().OpenACC().ActOnIntExpr(DK, CK, Loc, ER.get());
}
bool Parser::ParseOpenACCClauseVarList(OpenACCClauseKind Kind) {
@@ -739,7 +745,7 @@ bool Parser::ParseOpenACCSizeExprList() {
/// [num:]int-expr
/// dim:int-expr
/// static:size-expr
-bool Parser::ParseOpenACCGangArg() {
+bool Parser::ParseOpenACCGangArg(SourceLocation GangLoc) {
if (isOpenACCSpecialToken(OpenACCSpecialTokenKind::Static, getCurToken()) &&
NextToken().is(tok::colon)) {
@@ -753,7 +759,9 @@ bool Parser::ParseOpenACCGangArg() {
NextToken().is(tok::colon)) {
ConsumeToken();
ConsumeToken();
- return ParseOpenACCIntExpr().isInvalid();
+ return ParseOpenACCIntExpr(OpenACCDirectiveKind::Invalid,
+ OpenACCClauseKind::Gang, GangLoc)
+ .isInvalid();
}
if (isOpenACCSpecialToken(OpenACCSpecialTokenKind::Num, getCurToken()) &&
@@ -763,11 +771,13 @@ bool Parser::ParseOpenACCGangArg() {
// Fallthrough to the 'int-expr' handling for when 'num' is omitted.
}
// This is just the 'num' case where 'num' is optional.
- return ParseOpenACCIntExpr().isInvalid();
+ return ParseOpenACCIntExpr(OpenACCDirectiveKind::Invalid,
+ OpenACCClauseKind::Gang, GangLoc)
+ .isInvalid();
}
-bool Parser::ParseOpenACCGangArgList() {
- if (ParseOpenACCGangArg()) {
+bool Parser::ParseOpenACCGangArgList(SourceLocation GangLoc) {
+ if (ParseOpenACCGangArg(GangLoc)) {
SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end,
Parser::StopBeforeMatch);
return false;
@@ -776,7 +786,7 @@ bool Parser::ParseOpenACCGangArgList() {
while (!getCurToken().isOneOf(tok::r_paren, tok::annot_pragma_openacc_end)) {
ExpectAndConsume(tok::comma);
- if (ParseOpenACCGangArg()) {
+ if (ParseOpenACCGangArg(GangLoc)) {
SkipUntil(tok::r_paren, tok::annot_pragma_openacc_end,
Parser::StopBeforeMatch);
return false;
@@ -941,11 +951,18 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
case OpenACCClauseKind::DeviceNum:
case OpenACCClauseKind::DefaultAsync:
case OpenACCClauseKind::VectorLength: {
- ExprResult IntExpr = ParseOpenACCIntExpr();
+ ExprResult IntExpr = ParseOpenACCIntExpr(OpenACCDirectiveKind::Invalid,
+ ClauseKind, ClauseLoc);
if (IntExpr.isInvalid()) {
Parens.skipToEnd();
return OpenACCCanContinue();
}
+
+ // TODO OpenACC: as we implement the 'rest' of the above, this 'if' should
+ // be removed leaving just the 'setIntExprDetails'.
+ if (ClauseKind == OpenACCClauseKind::NumWorkers)
+ ParsedClause.setIntExprDetails(IntExpr.get());
+
break;
}
case OpenACCClauseKind::DType:
@@ -998,7 +1015,8 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
? OpenACCSpecialTokenKind::Length
: OpenACCSpecialTokenKind::Num,
ClauseKind);
- ExprResult IntExpr = ParseOpenACCIntExpr();
+ ExprResult IntExpr = ParseOpenACCIntExpr(OpenACCDirectiveKind::Invalid,
+ ClauseKind, ClauseLoc);
if (IntExpr.isInvalid()) {
Parens.skipToEnd();
return OpenACCCanContinue();
@@ -1014,13 +1032,14 @@ Parser::OpenACCClauseParseResult Parser::ParseOpenACCClauseParams(
break;
}
case OpenACCClauseKind::Gang:
- if (ParseOpenACCGangArgList()) {
+ if (ParseOpenACCGangArgList(ClauseLoc)) {
Parens.skipToEnd();
return OpenACCCanContinue();
}
break;
case OpenACCClauseKind::Wait:
- if (ParseOpenACCWaitArgument()) {
+ if (ParseOpenACCWaitArgument(ClauseLoc,
+ /*IsDirective=*/false)) {
Parens.skipToEnd();
return OpenACCCanContinue();
}
@@ -1052,7 +1071,7 @@ ExprResult Parser::ParseOpenACCAsyncArgument() {
/// In this section and throughout the specification, the term wait-argument
/// means:
/// [ devnum : int-expr : ] [ queues : ] async-argument-list
-bool Parser::ParseOpenACCWaitArgument() {
+bool Parser::ParseOpenACCWaitArgument(SourceLocation Loc, bool IsDirective) {
// [devnum : int-expr : ]
if (isOpenACCSpecialToken(OpenACCSpecialTokenKind::DevNum, Tok) &&
NextToken().is(tok::colon)) {
@@ -1061,7 +1080,11 @@ bool Parser::ParseOpenACCWaitArgument() {
// Consume colon.
ConsumeToken();
- ExprResult IntExpr = ParseOpenACCIntExpr();
+ ExprResult IntExpr = ParseOpenACCIntExpr(
+ IsDirective ? OpenACCDirectiveKind::Wait
+ : OpenACCDirectiveKind::Invalid,
+ IsDirective ? OpenACCClauseKind::Invalid : OpenACCClauseKind::Wait,
+ Loc);
if (IntExpr.isInvalid())
return true;
@@ -1245,7 +1268,7 @@ Parser::OpenACCDirectiveParseInfo Parser::ParseOpenACCDirective() {
break;
case OpenACCDirectiveKind::Wait:
// OpenACC has an optional paren-wrapped 'wait-argument'.
- if (ParseOpenACCWaitArgument())
+ if (ParseOpenACCWaitArgument(StartLoc, /*IsDirective=*/true))
T.skipToEnd();
else
T.consumeClose();
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 59f65eaf47a6da..39cde677ecc87a 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -14,6 +14,7 @@
#include "clang/Sema/SemaOpenACC.h"
#include "clang/AST/StmtOpenACC.h"
#include "clang/Basic/DiagnosticSema.h"
+#include "clang/Basic/OpenACCKinds.h"
#include "clang/Sema/Sema.h"
#include "llvm/Support/Casting.h"
@@ -90,6 +91,16 @@ bool doesClauseApplyToDirective(OpenACCDirectiveKind DirectiveKind,
default:
return false;
}
+ case OpenACCClauseKind::NumWorkers:
+ switch (DirectiveKind) {
+ case OpenACCDirectiveKind::Parallel:
+ case OpenACCDirectiveKind::Kernels:
+ case OpenACCDirectiveKind::ParallelLoop:
+ case OpenACCDirectiveKind::KernelsLoop:
+ return true;
+ default:
+ return false;
+ }
default:
// Do nothing so we can go to the 'unimplemented' diagnostic instead.
return true;
@@ -218,6 +229,25 @@ SemaOpenACC::ActOnClause(ArrayRef<const OpenACCClause *> ExistingClauses,
getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(),
Clause.getConditionExpr(), Clause.getEndLoc());
}
+ case OpenACCClauseKind::NumWorkers: {
+ // 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 NumWorkers");
+ return OpenACCNumWorkersClause::Create(
+ getASTContext(), Clause.getBeginLoc(), Clause.getLParenLoc(),
+ Clause.getIntExprs()[0], Clause.getEndLoc());
+ }
default:
break;
}
@@ -248,6 +278,96 @@ void SemaOpenACC::ActOnConstruct(OpenACCDirectiveKind K,
}
}
+ExprResult SemaOpenACC::ActOnIntExpr(OpenACCDirectiveKind DK,
+ OpenACCClauseKind CK, SourceLocation Loc,
+ Expr *IntExpr) {
+
+ assert(((DK != OpenACCDirectiveKind::Invalid &&
+ CK == OpenACCClauseKind::Invalid) ||
+ (DK == OpenACCDirectiveKind::Invalid &&
+ CK != OpenACCClauseKind::Invalid)) &&
+ "Only one of directive or clause kind should be provided");
+
+ class IntExprConverter : public Sema::ICEConvertDiagnoser {
+ OpenACCDirectiveKind DirectiveKind;
+ OpenACCClauseKind ClauseKind;
+ Expr *IntExpr;
+
+ public:
+ IntExprConverter(OpenACCDirectiveKind DK, OpenACCClauseKind CK,
+ Expr *IntExpr)
+ : ICEConvertDiagnoser(/*AllowScopedEnumerations=*/false,
+ /*Suppress=*/false,
+ /*SuppressConversion=*/true),
+ DirectiveKind(DK), ClauseKind(CK), IntExpr(IntExpr) {}
+
+ bool match(QualType T) override {
+ // OpenACC spec just calls this 'integer expression' as having an
+ // 'integer type', so fall back on C99's 'integer type'.
+ return T->isIntegerType();
+ }
+ SemaBase::SemaDiagnosticBuilder diagnoseNotInt(Sema &S, SourceLocation Loc,
+ QualType T) override {
+ if (ClauseKind != OpenACCClauseKind::Invalid)
+ return S.Diag(Loc, diag::err_acc_int_expr_requires_integer) <<
+ /*Clause=*/0 << ClauseKind << T;
+
+ return S.Diag(Loc, diag::err_acc_int_expr_requires_integer) <<
+ /*Directive=*/1 << DirectiveKind << T;
+ }
+
+ SemaBase::SemaDiagnosticBuilder
+ diagnoseIncomplete(Sema &S, SourceLocation Loc, QualType T) override {
+ return S.Diag(Loc, diag::err_acc_int_expr_incomplete_class_type)
+ << T << IntExpr->getSourceRange();
+ }
+
+ SemaBase::SemaDiagnosticBuilder
+ diagnoseExplicitConv(Sema &S, SourceLocation Loc, QualType T,
+ QualType ConvTy) override {
+ return S.Diag(Loc, diag::err_acc_int_expr_explicit_conversion)
+ << T << ConvTy;
+ }
+
+ SemaBase::SemaDiagnosticBuilder noteExplicitConv(Sema &S,
+ CXXConversionDecl *Conv,
+ QualType ConvTy) override {
+ return S.Diag(Conv->getLocation(), diag::note_acc_int_expr_conversion)
+ << ConvTy->isEnumeralType() << ConvTy;
+ }
+
+ SemaBase::SemaDiagnosticBuilder
+ diagnoseAmbiguous(Sema &S, SourceLocation Loc, QualType T) override {
+ return S.Diag(Loc, diag::err_acc_int_expr_multiple_conversions) << T;
+ }
+
+ SemaBase::SemaDiagnosticBuilder
+ noteAmbiguous(Sema &S, CXXConversionDecl *Conv, QualType ConvTy) override {
+ return S.Diag(Conv->getLocation(), diag::note_acc_int_expr_conversion)
+ << ConvTy->isEnumeralType() << ConvTy;
+ }
+
+ SemaBase::SemaDiagnosticBuilder
+ diagnoseConversion(Sema &S, SourceLocation Loc, QualType T,
+ QualType ConvTy) override {
+ llvm_unreachable("conversion functions are permitted");
+ }
+ } IntExprDiagnoser(DK, CK, IntExpr);
+
+ ExprResult IntExprResult = SemaRef.PerformContextualImplicitConversion(
+ Loc, IntExpr, IntExprDiagnoser);
+ if (IntExprResult.isInvalid())
+ return ExprError();
+
+ IntExpr = IntExprResult.get();
+ if (!IntExpr->isTypeDependent() && !IntExpr->getType()->isIntegerType())
+ return ExprError();
+
+ // TODO OpenACC: Do we want to perform usual unary conversions here? When
+ // doing codegen we might find that is necessary, but skip it for now.
+ return IntExpr;
+}
+
bool SemaOpenACC::ActOnStartStmtDirective(OpenACCDirectiveKind K,
SourceLocation StartLoc) {
return diagnoseConstructAppertainment(*this, K, StartLoc, /*IsStmt=*/true);
diff --git a/clang/lib/Sema/TreeTransform.h b/clang/lib/Sema/TreeTransform.h
index eb05783a6219dc..34b52a346a10d1 100644
--- a/clang/lib/Sema/TreeTransform.h
+++ b/clang/lib/Sema/TreeTransform.h
@@ -11158,6 +11158,29 @@ void OpenACCClauseTransform<Derived>::VisitSelfClause(
ParsedClause.getLParenLoc(), ParsedClause.getConditionExpr(),
ParsedClause.getEndLoc());
}
+
+template <typename Derived>
+void OpenACCClauseTransform<Derived>::VisitNumWorkersClause(
+ const OpenACCNumWorkersClause &C) {
+ Expr *IntExpr = const_cast<Expr *>(C.getIntExpr());
+ assert(IntExpr && "num_workers 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 = OpenACCNumWorkersClause::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 b28df03b4a95e9..6fff829cefb845 100644
--- a/clang/lib/Serialization/ASTReader.cpp
+++ b/clang/lib/Serialization/ASTReader.cpp
@@ -11786,6 +11786,12 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
return OpenACCSelfClause::Create(getContext(), BeginLoc, LParenLoc,
CondExpr, EndLoc);
}
+ case OpenACCClauseKind::NumWorkers: {
+ SourceLocation LParenLoc = readSourceLocation();
+ Expr *IntExpr = readSubExpr();
+ return OpenACCNumWorkersClause::Create(getContext(), BeginLoc, LParenLoc,
+ IntExpr, EndLoc);
+ }
case OpenACCClauseKind::Finalize:
case OpenACCClauseKind::IfPresent:
case OpenACCClauseKind::Seq:
@@ -11816,7 +11822,6 @@ OpenACCClause *ASTRecordReader::readOpenACCClause() {
case OpenACCClauseKind::Bind:
case OpenACCClauseKind::VectorLength:
case OpenACCClauseKind::NumGangs:
- case OpenACCClauseKind::NumWorkers:
case OpenACCClauseKind::DeviceNum:
case OpenACCClauseKind::DefaultAsync:
case OpenACCClauseKind::DeviceType:
diff --git a/clang/lib/Serialization/ASTWriter.cpp b/clang/lib/Serialization/ASTWriter.cpp
index 2eb4adac53de2a..934b202ac38267 100644
--- a/clang/lib/Serialization/ASTWriter.cpp
+++ b/clang/lib/Serialization/ASTWriter.cpp
@@ -7657,6 +7657,12 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
AddStmt(const_cast<Expr *>(SC->getConditionExpr()));
return;
}
+ case OpenACCClauseKind::NumWorkers: {
+ const auto *NWC = cast<OpenACCNumWorkersClause>(C);
+ writeSourceLocation(NWC->getLParenLoc());
+ AddStmt(const_cast<Expr *>(NWC->getIntExpr()));
+ return;
+ }
case OpenACCClauseKind::Finalize:
case OpenACCClauseKind::IfPresent:
case OpenACCClauseKind::Seq:
@@ -7687,7 +7693,6 @@ void ASTRecordWriter::writeOpenACCClause(const OpenACCClause *C) {
case OpenACCClauseKind::Bind:
case OpenACCClauseKind::VectorLength:
case OpenACCClauseKind::NumGangs:
- case OpenACCClauseKind::NumWorkers:
case OpenACCClauseKind::DeviceNum:
case OpenACCClauseKind::DefaultAsync:
case OpenACCClauseKind::DeviceType:
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index 4462f0df540f2d..bd5f459eae074f 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -943,11 +943,9 @@ void IntExprParsing() {
#pragma acc parallel num_workers(5, 4)
{}
- // expected-warning at +1{{OpenACC clause 'num_workers' not yet implemented, clause ignored}}
#pragma acc parallel num_workers(5)
{}
- // expected-warning at +1{{OpenACC clause 'num_workers' not yet implemented, clause ignored}}
#pragma acc parallel num_workers(returns_int())
{}
diff --git a/clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp b/clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp
new file mode 100644
index 00000000000000..e3664460a6e5e3
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-intexpr-clause-ast.cpp
@@ -0,0 +1,255 @@
+// RUN: %clang_cc1 %s -fopenacc -ast-dump | FileCheck %s
+
+// Test this with PCH.
+// RUN: %clang_cc1 %s -fopenacc -emit-pch -o %t %s
+// RUN: %clang_cc1 %s -fopenacc -include-pch %t -ast-dump-all | FileCheck %s
+
+#ifndef PCH_HELPER
+#define PCH_HELPER
+
+int some_int();
+short some_short();
+long some_long();
+enum E{};
+E some_enum();
+
+struct CorrectConvert {
+ operator int();
+} Convert;
+
+
+void NormalUses() {
+ // CHECK: FunctionDecl{{.*}}NormalUses
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel num_workers(some_int())
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CallExpr{{.*}}'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}}'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'int ()' lvalue Function{{.*}} 'some_int' 'int ()'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc kernels num_workers(some_short())
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers 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
+
+#pragma acc parallel num_workers(some_long())
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CallExpr{{.*}}'long'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}}'long (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'long ()' lvalue Function{{.*}} 'some_long' 'long ()'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel num_workers(some_enum())
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CallExpr{{.*}}'E'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}}'E (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}}'E ()' lvalue Function{{.*}} 'some_enum' 'E ()'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc kernels num_workers(Convert)
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <UserDefinedConversion>
+ // CHECK-NEXT: CXXMemberCallExpr{{.*}}'int'
+ // CHECK-NEXT: MemberExpr{{.*}} '<bound member function type>' .operator int
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'struct CorrectConvert':'CorrectConvert' lvalue Var
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+}
+
+template<typename T, typename U>
+void TemplUses(T t, U u) {
+ // CHECK-NEXT: FunctionTemplateDecl
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 0 T
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}}typename depth 0 index 1 U
+ // CHECK-NEXT: FunctionDecl{{.*}} TemplUses 'void (T, U)'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced t 'T'
+ // CHECK-NEXT: ParmVarDecl{{.*}} referenced u 'U'
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel num_workers(t)
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T' lvalue ParmVar{{.*}} 't' 'T'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc kernels num_workers(u)
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'U' lvalue ParmVar{{.*}} 'u' 'U'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel num_workers(U::value)
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: DependentScopeDeclRefExpr{{.*}} '<dependent type>' lvalue
+ // CHECK-NEXT: NestedNameSpecifier TypeSpec 'U'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc kernels num_workers(T{})
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CXXUnresolvedConstructExpr{{.*}} 'T' 'T' list
+ // CHECK-NEXT: InitListExpr{{.*}} 'void'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel num_workers(U{})
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CXXUnresolvedConstructExpr{{.*}} 'U' 'U' list
+ // CHECK-NEXT: InitListExpr{{.*}} 'void'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc kernels num_workers(typename U::IntTy{})
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CXXUnresolvedConstructExpr{{.*}} 'typename U::IntTy' 'typename U::IntTy' list
+ // CHECK-NEXT: InitListExpr{{.*}} 'void'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+#pragma acc parallel num_workers(typename U::ShortTy{})
+ while(true){}
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CXXUnresolvedConstructExpr{{.*}} 'typename U::ShortTy' 'typename U::ShortTy' list
+ // CHECK-NEXT: InitListExpr{{.*}} 'void'
+ // 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'
+ // CHECK-NEXT: RecordType{{.*}} 'CorrectConvert'
+ // CHECK-NEXT: CXXRecord{{.*}} 'CorrectConvert'
+ // CHECK-NEXT: TemplateArgument type 'HasInt'
+ // CHECK-NEXT: RecordType{{.*}} 'HasInt'
+ // CHECK-NEXT: CXXRecord{{.*}} 'HasInt'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used t 'CorrectConvert'
+ // CHECK-NEXT: ParmVarDecl{{.*}} used u 'HasInt'
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <UserDefinedConversion>
+ // CHECK-NEXT: CXXMemberCallExpr{{.*}}'int'
+ // CHECK-NEXT: MemberExpr{{.*}} '<bound member function type>' .operator int
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'CorrectConvert' lvalue ParmVar
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers 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: num_workers 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
+
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <UserDefinedConversion>
+ // CHECK-NEXT: CXXMemberCallExpr{{.*}}'int'
+ // CHECK-NEXT: MemberExpr{{.*}} '<bound member function type>' .operator int
+ // CHECK-NEXT: MaterializeTemporaryExpr{{.*}} 'CorrectConvert' lvalue
+ // CHECK-NEXT: CXXFunctionalCastExpr{{.*}} 'CorrectConvert' functional cast to struct CorrectConvert <NoOp>
+ // CHECK-NEXT: InitListExpr{{.*}}'CorrectConvert'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: ExprWithCleanups
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'char' <UserDefinedConversion>
+ // CHECK-NEXT: CXXMemberCallExpr{{.*}}'char'
+ // CHECK-NEXT: MemberExpr{{.*}} '<bound member function type>' .operator char
+ // CHECK-NEXT: MaterializeTemporaryExpr{{.*}} 'HasInt' lvalue
+ // CHECK-NEXT: CXXFunctionalCastExpr{{.*}} 'HasInt' functional cast to struct HasInt <NoOp>
+ // CHECK-NEXT: InitListExpr{{.*}}'HasInt'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: ExprWithCleanups
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}kernels
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CXXFunctionalCastExpr{{.*}} 'typename HasInt::IntTy':'int' functional cast to typename struct HasInt::IntTy <NoOp>
+ // CHECK-NEXT: InitListExpr{{.*}}'typename HasInt::IntTy':'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}}parallel
+ // CHECK-NEXT: num_workers clause
+ // CHECK-NEXT: CXXFunctionalCastExpr{{.*}} 'typename HasInt::ShortTy':'short' functional cast to typename struct HasInt::ShortTy <NoOp>
+ // CHECK-NEXT: InitListExpr{{.*}}'typename HasInt::ShortTy':'short'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: CompoundStmt
+}
+
+struct HasInt {
+ using IntTy = int;
+ using ShortTy = short;
+ static constexpr int value = 1;
+
+ operator char();
+};
+
+void Inst() {
+ TemplUses<CorrectConvert, HasInt>({}, {});
+}
+#endif // PCH_HELPER
diff --git a/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c b/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c
new file mode 100644
index 00000000000000..19e247a2f810ae
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c
@@ -0,0 +1,33 @@
+// RUN: %clang_cc1 %s -fopenacc -verify
+
+short getS();
+
+void Test() {
+#pragma acc parallel num_workers(1)
+ while(1);
+#pragma acc kernels num_workers(1)
+ while(1);
+
+ // expected-error at +1{{OpenACC 'num_workers' clause is not valid on 'serial' directive}}
+#pragma acc serial num_workers(1)
+ while(1);
+
+ struct NotConvertible{} NC;
+ // expected-error at +1{{OpenACC clause 'num_workers' requires expression of integer type ('struct NotConvertible' invalid)}}
+#pragma acc parallel num_workers(NC)
+ while(1);
+
+#pragma acc kernels num_workers(getS())
+ while(1);
+
+ struct Incomplete *SomeIncomplete;
+
+ // expected-error at +1{{OpenACC clause 'num_workers' requires expression of integer type ('struct Incomplete' invalid)}}
+#pragma acc kernels num_workers(*SomeIncomplete)
+ while(1);
+
+ enum E{A} SomeE;
+
+#pragma acc kernels num_workers(SomeE)
+ while(1);
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-num_workers-clause.cpp b/clang/test/SemaOpenACC/compute-construct-num_workers-clause.cpp
new file mode 100644
index 00000000000000..9449b77d092f4a
--- /dev/null
+++ b/clang/test/SemaOpenACC/compute-construct-num_workers-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 num_workers(1)
+ while(1);
+#pragma acc kernels num_workers(1)
+ while(1);
+
+ // expected-error at +1{{OpenACC clause 'num_workers' requires expression of integer type ('struct NotConvertible' invalid}}
+#pragma acc parallel num_workers(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 num_workers(*SomeIncomplete)
+ while(1);
+
+#pragma acc parallel num_workers(SomeE)
+ while(1);
+
+ // expected-error at +1{{OpenACC clause 'num_workers' requires expression of integer type ('enum E2' invalid}}
+#pragma acc kernels num_workers(SomeE2)
+ while(1);
+
+#pragma acc parallel num_workers(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 num_workers(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 num_workers(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 num_workers(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 num_workers(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 num_workers(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 num_workers(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 num_workers(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 num_workers(T::EXValue)
+ while (1);
+
+#pragma acc parallel num_workers(HasInt::value)
+ while (1);
+
+#pragma acc kernels num_workers(T::value)
+ while (1);
+
+#pragma acc parallel num_workers(HasInt::IntTy{})
+ while (1);
+
+#pragma acc kernels num_workers(typename T::ShortTy{})
+ while (1);
+
+#pragma acc parallel num_workers(HasInt::IntTy{})
+ while (1);
+
+#pragma acc kernels num_workers(typename T::ShortTy{})
+ while (1);
+
+ HasInt HI{};
+ T MyT{};
+
+#pragma acc parallel num_workers(HI)
+ while (1);
+
+#pragma acc kernels num_workers(MyT)
+ while (1);
+}
+
+void Inst() {
+ TestInst<HasInt>(); // #INST
+}
diff --git a/clang/tools/libclang/CIndex.cpp b/clang/tools/libclang/CIndex.cpp
index 2ef599d2cd26fa..011bec32bab319 100644
--- a/clang/tools/libclang/CIndex.cpp
+++ b/clang/tools/libclang/CIndex.cpp
@@ -2795,6 +2795,10 @@ void OpenACCClauseEnqueue::VisitSelfClause(const OpenACCSelfClause &C) {
if (C.hasConditionExpr())
Visitor.AddStmt(C.getConditionExpr());
}
+void OpenACCClauseEnqueue::VisitNumWorkersClause(
+ const OpenACCNumWorkersClause &C) {
+ Visitor.AddStmt(C.getIntExpr());
+}
} // namespace
void EnqueueVisitor::EnqueueChildren(const OpenACCClause *C) {
More information about the cfe-commits
mailing list