[clang] [OpenACC] Implement SubArray Parsing/Sema (PR #90796)
Erich Keane via cfe-commits
cfe-commits at lists.llvm.org
Thu May 2 06:23:17 PDT 2024
https://github.com/erichkeane updated https://github.com/llvm/llvm-project/pull/90796
>From ac69522bf15c7c169c4cdb9d3d1547c5c0962193 Mon Sep 17 00:00:00 2001
From: erichkeane <ekeane at nvidia.com>
Date: Tue, 30 Apr 2024 11:42:35 -0700
Subject: [PATCH 1/2] [OpenACC} Implement SubArray Parsing/Sema
This implementation takes quite a bit from the OMP implementation of
array sections, but only has to enforce the rules as applicable to
OpenACC. Additionally, it does its best to create an AST node (with the
assistance of RecoveryExprs) with as much checking done as soon as
possible in the case of instantiations.
---
.../clang/Basic/DiagnosticSemaKinds.td | 24 +-
clang/lib/Parse/ParseExpr.cpp | 3 +-
clang/lib/Sema/SemaOpenACC.cpp | 232 ++++++-
.../ParserOpenACC/parse-cache-construct.c | 3 +-
.../ParserOpenACC/parse-cache-construct.cpp | 8 +-
clang/test/ParserOpenACC/parse-clauses.c | 2 -
clang/test/ParserOpenACC/parse-sub-array.cpp | 89 +++
.../compute-construct-private-clause.c | 26 +-
.../compute-construct-private-clause.cpp | 3 +-
.../compute-construct-varlist-ast.cpp | 9 +
clang/test/SemaOpenACC/sub-array-ast.cpp | 566 ++++++++++++++++++
clang/test/SemaOpenACC/sub-array.cpp | 208 +++++++
12 files changed, 1131 insertions(+), 42 deletions(-)
create mode 100644 clang/test/ParserOpenACC/parse-sub-array.cpp
create mode 100644 clang/test/SemaOpenACC/sub-array-ast.cpp
create mode 100644 clang/test/SemaOpenACC/sub-array.cpp
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index 4b074b853bfe65..fbb107679d94ba 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -12291,8 +12291,8 @@ def warn_acc_if_self_conflict
"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)">;
+ : Error<"OpenACC %select{clause '%1'|directive '%2'|sub-array bound}0 "
+ "requires expression of integer type (%3 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
@@ -12310,4 +12310,24 @@ def err_acc_num_gangs_num_args
def err_acc_not_a_var_ref
: Error<"OpenACC variable is not a valid variable name, sub-array, array "
"element, or composite variable member">;
+def err_acc_typecheck_subarray_value
+ : Error<"OpenACC sub-array subscripted value is not an array or pointer">;
+def err_acc_subarray_function_type
+ : Error<"OpenACC sub-array cannot be of function type %0">;
+def err_acc_subarray_incomplete_type
+ : Error<"OpenACC sub-array base is of incomplete type %0">;
+def err_acc_subarray_no_length
+ : Error<"OpenACC sub-array length is unspecified and cannot be inferred "
+ "because the subscripted value is %select{not an array|an array of "
+ "unknown bound}0">;
+def err_acc_subarray_negative
+ : Error<"OpenACC sub-array %select{lower bound|length}0 evaluated to "
+ "negative value %1">;
+def err_acc_subarray_out_of_range
+ : Error<"OpenACC sub-array %select{lower bound|length}0 evaluated to a "
+ "value (%1) that would be out of the range of the subscripted "
+ "array size of %2">;
+def err_acc_subarray_base_plus_length_out_of_range
+ : Error<"OpenACC sub-array specified range [%0:%1] would be out of the "
+ "range of the subscripted array size of %2">;
} // end of sema component.
diff --git a/clang/lib/Parse/ParseExpr.cpp b/clang/lib/Parse/ParseExpr.cpp
index 7d6febb04a82c4..5f5f9a79c8c408 100644
--- a/clang/lib/Parse/ParseExpr.cpp
+++ b/clang/lib/Parse/ParseExpr.cpp
@@ -2039,7 +2039,8 @@ Parser::ParsePostfixExpressionSuffix(ExprResult LHS) {
if (Tok.is(tok::colon)) {
// Consume ':'
ColonLocFirst = ConsumeToken();
- Length = Actions.CorrectDelayedTyposInExpr(ParseExpression());
+ if (Tok.isNot(tok::r_square))
+ Length = Actions.CorrectDelayedTyposInExpr(ParseExpression());
}
} else if (ArgExprs.size() <= 1 && getLangOpts().OpenMP) {
ColonProtectionRAIIObject RAII(*this);
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index 3ea81e0497c203..a203ece7ff3130 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -16,6 +16,7 @@
#include "clang/Basic/DiagnosticSema.h"
#include "clang/Basic/OpenACCKinds.h"
#include "clang/Sema/Sema.h"
+#include "llvm/ADT/StringExtras.h"
#include "llvm/Support/Casting.h"
using namespace clang;
@@ -367,7 +368,9 @@ ExprResult SemaOpenACC::ActOnIntExpr(OpenACCDirectiveKind DK,
assert(((DK != OpenACCDirectiveKind::Invalid &&
CK == OpenACCClauseKind::Invalid) ||
(DK == OpenACCDirectiveKind::Invalid &&
- CK != OpenACCClauseKind::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 {
@@ -375,6 +378,16 @@ ExprResult SemaOpenACC::ActOnIntExpr(OpenACCDirectiveKind DK,
OpenACCClauseKind ClauseKind;
Expr *IntExpr;
+ // gets the index into the diagnostics so we can use this for clauses,
+ // directives, and sub array.s
+ unsigned getDiagKind() const {
+ if (ClauseKind != OpenACCClauseKind::Invalid)
+ return 0;
+ if (DirectiveKind != OpenACCDirectiveKind::Invalid)
+ return 1;
+ return 2;
+ }
+
public:
IntExprConverter(OpenACCDirectiveKind DK, OpenACCClauseKind CK,
Expr *IntExpr)
@@ -390,12 +403,8 @@ ExprResult SemaOpenACC::ActOnIntExpr(OpenACCDirectiveKind DK,
}
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;
+ return S.Diag(Loc, diag::err_acc_int_expr_requires_integer)
+ << getDiagKind() << ClauseKind << DirectiveKind << T;
}
SemaBase::SemaDiagnosticBuilder
@@ -503,12 +512,211 @@ ExprResult SemaOpenACC::ActOnArraySectionExpr(Expr *Base, SourceLocation LBLoc,
SourceLocation RBLoc) {
ASTContext &Context = getASTContext();
- // TODO OpenACC: We likely have to reproduce a lot of the same logic from the
- // OMP version of this, but at the moment we don't have a good way to test it,
- // so for now we'll just create the node.
+ // Handle placeholders.
+ if (Base->hasPlaceholderType() &&
+ !Base->hasPlaceholderType(BuiltinType::ArraySection)) {
+ ExprResult Result = SemaRef.CheckPlaceholderExpr(Base);
+ if (Result.isInvalid())
+ return ExprError();
+ Base = Result.get();
+ }
+ if (LowerBound && LowerBound->getType()->isNonOverloadPlaceholderType()) {
+ ExprResult Result = SemaRef.CheckPlaceholderExpr(LowerBound);
+ if (Result.isInvalid())
+ return ExprError();
+ Result = SemaRef.DefaultLvalueConversion(Result.get());
+ if (Result.isInvalid())
+ return ExprError();
+ LowerBound = Result.get();
+ }
+ if (Length && Length->getType()->isNonOverloadPlaceholderType()) {
+ ExprResult Result = SemaRef.CheckPlaceholderExpr(Length);
+ if (Result.isInvalid())
+ return ExprError();
+ Result = SemaRef.DefaultLvalueConversion(Result.get());
+ if (Result.isInvalid())
+ return ExprError();
+ Length = Result.get();
+ }
+
+ // Check the 'base' value, it must be an array or pointer type, and not to/of
+ // a function type.
+ QualType OriginalBaseTy = ArraySectionExpr::getBaseOriginalType(Base);
+ QualType ResultTy;
+ if (!Base->isTypeDependent()) {
+ if (OriginalBaseTy->isAnyPointerType()) {
+ ResultTy = OriginalBaseTy->getPointeeType();
+ } else if (OriginalBaseTy->isArrayType()) {
+ ResultTy = OriginalBaseTy->getAsArrayTypeUnsafe()->getElementType();
+ } else {
+ return ExprError(
+ Diag(Base->getExprLoc(), diag::err_acc_typecheck_subarray_value)
+ << Base->getSourceRange());
+ }
+
+ if (ResultTy->isFunctionType()) {
+ Diag(Base->getExprLoc(), diag::err_acc_subarray_function_type)
+ << ResultTy << Base->getSourceRange();
+ return ExprError();
+ }
+
+ if (SemaRef.RequireCompleteType(Base->getExprLoc(), ResultTy,
+ diag::err_acc_subarray_incomplete_type,
+ Base))
+ return ExprError();
+
+ if (!Base->hasPlaceholderType(BuiltinType::ArraySection)) {
+ ExprResult Result = SemaRef.DefaultFunctionArrayLvalueConversion(Base);
+ if (Result.isInvalid())
+ return ExprError();
+ Base = Result.get();
+ }
+ }
+
+ auto GetRecovery = [&](Expr *E, QualType Ty) {
+ ExprResult Recovery =
+ SemaRef.CreateRecoveryExpr(E->getBeginLoc(), E->getEndLoc(), E, Ty);
+ return Recovery.isUsable() ? Recovery.get() : nullptr;
+ };
+
+ // Ensure both of the expressions are int-exprs.
+ if (LowerBound && !LowerBound->isTypeDependent()) {
+ ExprResult LBRes =
+ ActOnIntExpr(OpenACCDirectiveKind::Invalid, OpenACCClauseKind::Invalid,
+ LowerBound->getExprLoc(), LowerBound);
+
+ if (LBRes.isUsable())
+ LBRes = SemaRef.DefaultLvalueConversion(LBRes.get());
+ LowerBound =
+ LBRes.isUsable() ? LBRes.get() : GetRecovery(LowerBound, Context.IntTy);
+ }
+
+ if (Length && !Length->isTypeDependent()) {
+ ExprResult LenRes =
+ ActOnIntExpr(OpenACCDirectiveKind::Invalid, OpenACCClauseKind::Invalid,
+ Length->getExprLoc(), Length);
+
+ if (LenRes.isUsable())
+ LenRes = SemaRef.DefaultLvalueConversion(LenRes.get());
+ Length =
+ LenRes.isUsable() ? LenRes.get() : GetRecovery(Length, Context.IntTy);
+ }
+
+ // Length is required if the base type is not an array of known bounds.
+ if (!Length && (OriginalBaseTy.isNull() ||
+ (!OriginalBaseTy->isDependentType() &&
+ !OriginalBaseTy->isConstantArrayType() &&
+ !OriginalBaseTy->isDependentSizedArrayType()))) {
+ bool IsArray = !OriginalBaseTy.isNull() && OriginalBaseTy->isArrayType();
+ Diag(ColonLoc, diag::err_acc_subarray_no_length) << IsArray;
+ // Fill in a dummy 'length' so that when we instantiate this we don't
+ // double-diagnose here.
+ ExprResult Recovery = SemaRef.CreateRecoveryExpr(
+ ColonLoc, SourceLocation(), ArrayRef<Expr *>{std::nullopt},
+ Context.IntTy);
+ Length = Recovery.isUsable() ? Recovery.get() : nullptr;
+ }
+
+ // Check the values of each of the arguments, they cannot be negative(we
+ // assume), and if the array bound is known, must be within range. As we do
+ // so, do our best to continue with evaluation, we can set the
+ // value/expression to nullptr/nullopt if they are invalid, and treat them as
+ // not present for the rest of evaluation.
+
+ // We don't have to check for dependence, because the dependent size is
+ // represented as a different AST node.
+ std::optional<llvm::APSInt> BaseSize;
+ if (!OriginalBaseTy.isNull() && OriginalBaseTy->isConstantArrayType()) {
+ const auto *ArrayTy = Context.getAsConstantArrayType(OriginalBaseTy);
+ BaseSize = ArrayTy->getSize();
+ }
+
+ auto GetBoundValue = [&](Expr *E) -> std::optional<llvm::APSInt> {
+ if (!E || E->isInstantiationDependent())
+ return std::nullopt;
+
+ Expr::EvalResult Res;
+ if (!E->EvaluateAsInt(Res, Context))
+ return std::nullopt;
+ return Res.Val.getInt();
+ };
+
+ std::optional<llvm::APSInt> LowerBoundValue = GetBoundValue(LowerBound);
+ std::optional<llvm::APSInt> LengthValue = GetBoundValue(Length);
+
+ // Check lower bound for negative or out of range.
+ if (LowerBoundValue.has_value()) {
+ if (LowerBoundValue->isNegative()) {
+ Diag(LowerBound->getExprLoc(), diag::err_acc_subarray_negative)
+ << /*LowerBound=*/0 << toString(*LowerBoundValue, /*Radix=*/10);
+ LowerBoundValue.reset();
+ LowerBound = GetRecovery(LowerBound, LowerBound->getType());
+ } else if (BaseSize.has_value() &&
+ llvm::APSInt::compareValues(*LowerBoundValue, *BaseSize) >= 0) {
+ // Lower bound (start index) must be less than the size of the array.
+ Diag(LowerBound->getExprLoc(), diag::err_acc_subarray_out_of_range)
+ << /*LowerBound=*/0 << toString(*LowerBoundValue, /*Radix=*/10)
+ << toString(*BaseSize, /*Radix=*/10);
+ LowerBoundValue.reset();
+ LowerBound = GetRecovery(LowerBound, LowerBound->getType());
+ }
+ }
+
+ // Check length for negative or out of range.
+ if (LengthValue.has_value()) {
+ if (LengthValue->isNegative()) {
+ Diag(Length->getExprLoc(), diag::err_acc_subarray_negative)
+ << /*Length=*/1 << toString(*LengthValue, /*Radix=*/10);
+ LengthValue.reset();
+ Length = GetRecovery(Length, Length->getType());
+ } else if (BaseSize.has_value() &&
+ llvm::APSInt::compareValues(*LengthValue, *BaseSize) > 0) {
+ // Length must be lessthan or EQUAL to the size of the array.
+ Diag(Length->getExprLoc(), diag::err_acc_subarray_out_of_range)
+ << /*Length=*/1 << toString(*LengthValue, /*Radix=*/10)
+ << toString(*BaseSize, /*Radix=*/10);
+ LengthValue.reset();
+ Length = GetRecovery(Length, Length->getType());
+ }
+ }
+
+ // Adding two APSInts requires matching sign, so extract that here.
+ auto AddAPSInt = [](llvm::APSInt LHS, llvm::APSInt RHS) -> llvm::APSInt {
+ if (LHS.isSigned() == RHS.isSigned())
+ return LHS + RHS;
+
+ unsigned width = std::max(LHS.getBitWidth(), RHS.getBitWidth()) + 1;
+ return llvm::APSInt(LHS.sext(width) + RHS.sext(width), /*Signed=*/true);
+ };
+
+ // If we know all 3 values, we can diagnose that the total value would be out
+ // of range.
+ if (BaseSize.has_value() && LowerBoundValue.has_value() &&
+ LengthValue.has_value() &&
+ llvm::APSInt::compareValues(AddAPSInt(*LowerBoundValue, *LengthValue),
+ *BaseSize) > 0) {
+ Diag(Base->getExprLoc(),
+ diag::err_acc_subarray_base_plus_length_out_of_range)
+ << toString(*LowerBoundValue, /*Radix=*/10)
+ << toString(*LengthValue, /*Radix=*/10)
+ << toString(*BaseSize, /*Radix=*/10);
+
+ LowerBoundValue.reset();
+ LowerBound = GetRecovery(LowerBound, LowerBound->getType());
+ LengthValue.reset();
+ Length = GetRecovery(Length, Length->getType());
+ }
+
+ // If any part of the expression is dependent, return a dependent sub-array.
+ QualType ArrayExprTy = Context.ArraySectionTy;
+ if (Base->isTypeDependent() ||
+ (LowerBound && LowerBound->isInstantiationDependent()) ||
+ (Length && Length->isInstantiationDependent()))
+ ArrayExprTy = Context.DependentTy;
+
return new (Context)
- ArraySectionExpr(Base, LowerBound, Length, Context.ArraySectionTy,
- VK_LValue, OK_Ordinary, ColonLoc, RBLoc);
+ ArraySectionExpr(Base, LowerBound, Length, ArrayExprTy, VK_LValue,
+ OK_Ordinary, ColonLoc, RBLoc);
}
bool SemaOpenACC::ActOnStartStmtDirective(OpenACCDirectiveKind K,
diff --git a/clang/test/ParserOpenACC/parse-cache-construct.c b/clang/test/ParserOpenACC/parse-cache-construct.c
index de26fc2b277a6b..8937aa095d5eaa 100644
--- a/clang/test/ParserOpenACC/parse-cache-construct.c
+++ b/clang/test/ParserOpenACC/parse-cache-construct.c
@@ -134,9 +134,8 @@ void func() {
}
for (int i = 0; i < 10; ++i) {
- // expected-error at +2{{expected expression}}
// expected-warning at +1{{OpenACC construct 'cache' not yet implemented, pragma ignored}}
- #pragma acc cache(readonly:ArrayPtr[5:])
+ #pragma acc cache(readonly:ArrayPtr[5:1])
}
for (int i = 0; i < 10; ++i) {
diff --git a/clang/test/ParserOpenACC/parse-cache-construct.cpp b/clang/test/ParserOpenACC/parse-cache-construct.cpp
index f1c71e8b584786..374fe2697b63fc 100644
--- a/clang/test/ParserOpenACC/parse-cache-construct.cpp
+++ b/clang/test/ParserOpenACC/parse-cache-construct.cpp
@@ -74,12 +74,12 @@ void use() {
for (int i = 0; i < 10; ++i) {
// expected-error at +2{{OpenACC sub-array is not allowed here}}
// expected-warning at +1{{OpenACC construct 'cache' not yet implemented, pragma ignored}}
- #pragma acc cache(Arrs.MemArr[3:4].array[1:4])
+ #pragma acc cache(Arrs.MemArr[2:1].array[1:4])
}
for (int i = 0; i < 10; ++i) {
// expected-error at +2{{OpenACC sub-array is not allowed here}}
// expected-warning at +1{{OpenACC construct 'cache' not yet implemented, pragma ignored}}
- #pragma acc cache(Arrs.MemArr[3:4].array[4])
+ #pragma acc cache(Arrs.MemArr[2:1].array[4])
}
for (int i = 0; i < 10; ++i) {
// expected-error at +3{{expected ']'}}
@@ -88,7 +88,7 @@ void use() {
#pragma acc cache(Arrs.MemArr[3:4:].array[4])
}
for (int i = 0; i < 10; ++i) {
- // expected-error at +2{{expected expression}}
+ // expected-error at +2{{OpenACC sub-array is not allowed here}}
// expected-warning at +1{{OpenACC construct 'cache' not yet implemented, pragma ignored}}
#pragma acc cache(Arrs.MemArr[:].array[4])
}
@@ -105,7 +105,7 @@ void use() {
#pragma acc cache(Arrs.MemArr[: :].array[4])
}
for (int i = 0; i < 10; ++i) {
- // expected-error at +2{{expected expression}}
+ // expected-error at +2{{OpenACC sub-array is not allowed here}}
// expected-warning at +1{{OpenACC construct 'cache' not yet implemented, pragma ignored}}
#pragma acc cache(Arrs.MemArr[3:].array[4])
}
diff --git a/clang/test/ParserOpenACC/parse-clauses.c b/clang/test/ParserOpenACC/parse-clauses.c
index 8a439a5ccd4bdc..fa43f42585fc2b 100644
--- a/clang/test/ParserOpenACC/parse-clauses.c
+++ b/clang/test/ParserOpenACC/parse-clauses.c
@@ -499,7 +499,6 @@ void VarListClauses() {
#pragma acc serial copy(HasMem.MemArr[1:3].array[1:2]), seq
for(;;){}
- // expected-error at +3{{expected expression}}
// expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[:]), seq
@@ -519,7 +518,6 @@ void VarListClauses() {
#pragma acc serial copy(HasMem.MemArr[: :]), seq
for(;;){}
- // expected-error at +3{{expected expression}}
// expected-warning at +2{{OpenACC clause 'copy' not yet implemented, clause ignored}}
// expected-warning at +1{{OpenACC clause 'seq' not yet implemented, clause ignored}}
#pragma acc serial copy(HasMem.MemArr[3:]), seq
diff --git a/clang/test/ParserOpenACC/parse-sub-array.cpp b/clang/test/ParserOpenACC/parse-sub-array.cpp
new file mode 100644
index 00000000000000..c0d3f89159e890
--- /dev/null
+++ b/clang/test/ParserOpenACC/parse-sub-array.cpp
@@ -0,0 +1,89 @@
+// RUN: %clang_cc1 %s -verify -fopenacc
+
+void Func(int i, int j) {
+ int array[5];
+#pragma acc parallel private(array[:])
+ while (true);
+#pragma acc parallel private(array[i:])
+ while (true);
+#pragma acc parallel private(array[:j])
+ while (true);
+#pragma acc parallel private(array[i:j])
+ while (true);
+#pragma acc parallel private(array[1:2])
+ while (true);
+
+ // expected-error at +1{{expected unqualified-id}}
+#pragma acc parallel private(array[::])
+ while (true);
+ // expected-error at +2{{expected ']'}}
+ // expected-note at +1{{to match this '['}}
+#pragma acc parallel private(array[1::])
+ while (true);
+ // expected-error at +2{{expected ']'}}
+ // expected-note at +1{{to match this '['}}
+#pragma acc parallel private(array[:2:])
+ while (true);
+ // expected-error at +3{{expected unqualified-id}}
+ // expected-error at +2{{expected ']'}}
+ // expected-note at +1{{to match this '['}}
+#pragma acc parallel private(array[::3])
+ while (true);
+ // expected-error at +2{{expected ']'}}
+ // expected-note at +1{{to match this '['}}
+#pragma acc parallel private(array[1:2:3])
+ while (true);
+}
+
+template<typename T, unsigned I, auto &IPtr>// #IPTR
+void TemplFunc() {
+ T array[I];
+ T array2[2*I];
+ T t; // #tDecl
+#pragma acc parallel private(array[:])
+ while (true);
+#pragma acc parallel private(array[t:])
+ while (true);
+#pragma acc parallel private(array[I-1:])
+ while (true);
+#pragma acc parallel private(array[IPtr:])
+ while (true);
+#pragma acc parallel private(array[:t])
+ while (true);
+#pragma acc parallel private(array[:I])
+ while (true);
+#pragma acc parallel private(array[:IPtr])
+ while (true);
+#pragma acc parallel private(array[t:t])
+ while (true);
+#pragma acc parallel private(array2[I:I])
+ while (true);
+#pragma acc parallel private(array[IPtr:IPtr])
+ while (true);
+
+ // expected-error at +1{{expected unqualified-id}}
+#pragma acc parallel private(array[::])
+ while (true);
+ // expected-error at +3{{'t' is not a class, namespace, or enumeration}}
+ // expected-note@#tDecl{{'t' declared here}}
+ // expected-error at +1{{expected unqualified-id}}
+#pragma acc parallel private(array[t::])
+ while (true);
+ // expected-error at +2{{expected ']'}}
+ // expected-note at +1{{to match this '['}}
+#pragma acc parallel private(array[:I:])
+ while (true);
+ // expected-error at +2{{no member named 'IPtr' in the global namespace}}
+ // expected-note@#IPTR{{'IPtr' declared here}}
+#pragma acc parallel private(array[::IPtr])
+ while (true);
+ // expected-error at +2{{expected ']'}}
+ // expected-note at +1{{to match this '['}}
+#pragma acc parallel private(array[IPtr:I:t])
+ while (true);
+}
+
+void use() {
+ static constexpr int SomeI = 1;
+ TemplFunc<int, 5, SomeI>();
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-private-clause.c b/clang/test/SemaOpenACC/compute-construct-private-clause.c
index 15775279fc8690..d2615c384cdb1c 100644
--- a/clang/test/SemaOpenACC/compute-construct-private-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-private-clause.c
@@ -12,12 +12,12 @@ typedef struct IsComplete {
int GlobalInt;
float GlobalArray[5];
-void *GlobalPointer;
+short *GlobalPointer;
Complete GlobalComposite;
-void uses(int IntParam, void *PointerParam, float ArrayParam[5], Complete CompositeParam) {
+void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete CompositeParam) {
int LocalInt;
- void *LocalPointer;
+ short *LocalPointer;
float LocalArray[5];
Complete LocalComposite;
@@ -35,17 +35,13 @@ void uses(int IntParam, void *PointerParam, float ArrayParam[5], Complete Compos
while(1);
#pragma acc parallel private(LocalArray)
while(1);
- // TODO OpenACC: Fix array sections, this should be allowed.
- // expected-error at +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 at +1{{expected expression}}
#pragma acc parallel private(LocalArray[2:])
while(1);
-#pragma acc parallel private(LocalArray[2:5])
+#pragma acc parallel private(LocalArray[2:1])
while(1);
#pragma acc parallel private(LocalArray[2])
while(1);
@@ -103,40 +99,36 @@ void uses(int IntParam, void *PointerParam, float ArrayParam[5], Complete Compos
#pragma acc parallel private(+GlobalInt)
while(1);
- // TODO OpenACC: Fix array sections, this should be allowed.
- // expected-error at +1{{expected expression}}
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
#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 at +1{{expected expression}}
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
#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 at +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 at +1{{expected expression}}
#pragma acc parallel private(ArrayParam[2:])
while(1);
+ // expected-error at +1{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
#pragma acc parallel private(ArrayParam[2:5])
while(1);
#pragma acc parallel private(ArrayParam[2])
while(1);
- // expected-error at +1{{OpenACC sub-array is not allowed here}}
+ // expected-error at +2{{OpenACC sub-array specified range [2:5] would be out of the range of the subscripted array size of 5}}
+ // expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
#pragma acc parallel private((float*)ArrayParam[2:5])
while(1);
// expected-error at +1{{OpenACC variable is not a valid variable name, sub-array, array element, or composite variable member}}
diff --git a/clang/test/SemaOpenACC/compute-construct-private-clause.cpp b/clang/test/SemaOpenACC/compute-construct-private-clause.cpp
index 4dd4e0d8029d67..a776b16f0feb27 100644
--- a/clang/test/SemaOpenACC/compute-construct-private-clause.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-private-clause.cpp
@@ -112,8 +112,7 @@ void TemplUses(T t, T (&arrayT)[I], V TemplComp) {
while(true);
#pragma acc parallel private(Pointer[:t])
while(true);
- // TODO OpenACC: When fixing sub-arrays, this should be permitted}}
- // expected-error at +1{{expected expression}}
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
#pragma acc parallel private(Pointer[1:])
while(true);
}
diff --git a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
index 341be3c58ebd21..aad4b9963c70db 100644
--- a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
+++ b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp
@@ -56,8 +56,11 @@ void NormalUses(float *PointerParam) {
// CHECK-NEXT: private clause
// CHECK-NEXT: DeclRefExpr{{.*}}'short[5]' lvalue Var{{.*}}'GlobalArray' 'short[5]'
// CHECK-NEXT: ArraySectionExpr
+ // 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: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int'
// CHECK-NEXT: WhileStmt
// CHECK-NEXT: CXXBoolLiteralExpr
@@ -241,8 +244,10 @@ void TemplUses(T t, U u, T*PointerParam) {
// CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
// CHECK-NEXT: private clause
// CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *'
// CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
// CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int'
// CHECK-NEXT: WhileStmt
// CHECK-NEXT: CXXBoolLiteralExpr
@@ -290,6 +295,7 @@ struct S {
// CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
// CHECK-NEXT: private clause
// CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
// CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
// CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
// CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
@@ -331,6 +337,7 @@ struct S {
// CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
// CHECK-NEXT: private clause
// CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
// CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
// CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
// CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
@@ -372,6 +379,7 @@ void S::foo() {
// CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
// CHECK-NEXT: private clause
// CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
// CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
// CHECK-NEXT: CXXThisExpr{{.*}} 'S *' implicit this
// CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
@@ -522,6 +530,7 @@ struct STempl {
// CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
// CHECK-NEXT: private clause
// CHECK-NEXT: ArraySectionExpr{{.*}}
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
// CHECK-NEXT: MemberExpr{{.*}} 'int[5]' lvalue ->ThisMemberArray
// CHECK-NEXT: CXXThisExpr{{.*}} 'STempl<int> *' implicit this
// CHECK-NEXT: IntegerLiteral{{.*}}'int' 1
diff --git a/clang/test/SemaOpenACC/sub-array-ast.cpp b/clang/test/SemaOpenACC/sub-array-ast.cpp
new file mode 100644
index 00000000000000..094976e1642752
--- /dev/null
+++ b/clang/test/SemaOpenACC/sub-array-ast.cpp
@@ -0,0 +1,566 @@
+// 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
+
+constexpr int returns_3() { return 3; }
+
+void Func(int i, int j) {
+ // CHECK: FunctionDecl{{.*}}Func
+ // CHECK-NEXT: ParmVarDecl{{.*}} i 'int'
+ // CHECK-NEXT: ParmVarDecl{{.*}} j 'int'
+ // CHECK-NEXT: CompoundStmt
+ int array[5];
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} array 'int[5]'
+ int VLA[i];
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} VLA 'int[i]'
+ int *ptr;
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} ptr 'int *'
+
+#pragma acc parallel private(array[returns_3():])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(array[:1])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(array[returns_3():1])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(array[i:j])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'j' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(VLA[:1])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[i]' lvalue Var{{.*}} 'VLA' 'int[i]'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(VLA[returns_3():1])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[i]' lvalue Var{{.*}} 'VLA' 'int[i]'
+ // CHECK-NEXT: CallExpr{{.*}}'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(VLA[i:j])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[i]' lvalue Var{{.*}} 'VLA' 'int[i]'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'j' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ptr[:1])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int *' lvalue Var{{.*}} 'ptr' 'int *'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ptr[returns_3():1])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int *' lvalue Var{{.*}} 'ptr' 'int *'
+ // CHECK-NEXT: CallExpr{{.*}}'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 1
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ptr[i:j])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int *' lvalue Var{{.*}} 'ptr' 'int *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'j' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+
+template<typename T, unsigned I, auto &CEArray>
+void Templ(int i){
+ // CHECK-NEXT: FunctionTemplateDecl{{.*}}Templ
+ // CHECK-NEXT: TemplateTypeParmDecl{{.*}} typename depth 0 index 0 T
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'auto &' depth 0 index 2 CEArray
+ // CHECK-NEXT: FunctionDecl{{.*}}Templ 'void (int)'
+ // CHECK-NEXT: ParmVarDecl{{.*}} i 'int'
+ // CHECK-NEXT: CompoundStmt
+ T array[I+2];
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} array 'T[I + 2]'
+ T VLA[i];
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} VLA 'T[i]'
+ T *ptr;
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} ptr 'T *'
+
+#pragma acc parallel private(array[returns_3():])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T[I + 2]' lvalue Var{{.*}} 'array' 'T[I + 2]'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}}'returns_3' 'int ()'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(array[:I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T[I + 2]' lvalue Var{{.*}} 'array' 'T[I + 2]'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(array[returns_3()-2:I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T[I + 2]' lvalue Var{{.*}} 'array' 'T[I + 2]'
+ // CHECK-NEXT: BinaryOperator{{.*}} 'int' '-'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 2
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(array[i:i])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T[I + 2]' lvalue Var{{.*}} 'array' 'T[I + 2]'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(VLA[:I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T[i]' lvalue Var{{.*}} 'VLA' 'T[i]'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(VLA[returns_3():I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T[i]' lvalue Var{{.*}} 'VLA' 'T[i]'
+ // CHECK-NEXT: CallExpr{{.*}}'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(VLA[i:i])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T[i]' lvalue Var{{.*}} 'VLA' 'T[i]'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ptr[:I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T *' lvalue Var{{.*}} 'ptr' 'T *'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ptr[returns_3():I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T *' lvalue Var{{.*}} 'ptr' 'T *'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(ptr[i:i])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'T *' lvalue Var{{.*}} 'ptr' 'T *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(CEArray[returns_3() - 2: I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'auto' lvalue NonTypeTemplateParm{{.*}} 'CEArray' 'auto &'
+ // CHECK-NEXT: BinaryOperator{{.*}} 'int' '-'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 2
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+#pragma acc parallel private(CEArray[: I])
+ while (true);
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'auto' lvalue NonTypeTemplateParm{{.*}} 'CEArray' 'auto &'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'unsigned int' NonTypeTemplateParm{{.*}} 'I' 'unsigned int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+
+ // Instantiation:
+ // CHECK-NEXT: FunctionDecl{{.*}} Templ 'void (int)' implicit_instantiation
+ // CHECK-NEXT: TemplateArgument{{.*}} 'int'
+ // CHECK-NEXT: BuiltinType{{.*}} 'int'
+ // CHECK-NEXT: TemplateArgument integral 3
+ // CHECK-NEXT: TemplateArgument decl
+ // CHECK-NEXT: Var{{.*}} 'CEArray' 'const int[5]'
+ // CHECK-NEXT: ParmVarDecl{{.*}} i 'int'
+ // CHECK-NEXT: CompoundStmt
+
+ // T array[I+2];
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} array 'int[5]'
+ // T VLA[i];
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} VLA 'int[i]'
+ // T *ptr;
+ // CHECK-NEXT: DeclStmt
+ // CHECK-NEXT: VarDecl{{.*}} ptr 'int *'
+
+//#pragma acc parallel private(array[returns_3():])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}}'returns_3' 'int ()'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(array[:I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(array[returns_3()-2:I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: BinaryOperator{{.*}} 'int' '-'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 2
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(array[i:i])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[5]' lvalue Var{{.*}} 'array' 'int[5]'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(VLA[:I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[i]' lvalue Var{{.*}} 'VLA' 'int[i]'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(VLA[returns_3():I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[i]' lvalue Var{{.*}} 'VLA' 'int[i]'
+ // CHECK-NEXT: CallExpr{{.*}}'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(VLA[i:i])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int[i]' lvalue Var{{.*}} 'VLA' 'int[i]'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(ptr[:I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int *' lvalue Var{{.*}} 'ptr' 'int *'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(ptr[returns_3():I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int *' lvalue Var{{.*}} 'ptr' 'int *'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(ptr[i:i])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int *' lvalue Var{{.*}} 'ptr' 'int *'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int' lvalue ParmVar{{.*}} 'i' 'int'
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(CEArray[returns_3() - 2: I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}}'const int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'const int[5]' lvalue
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'auto &' depth 0 index 2 CEArray
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'const int[5]' lvalue Var{{.*}}'CEArray' 'const int[5]'
+ // CHECK-NEXT: BinaryOperator{{.*}} 'int' '-'
+ // CHECK-NEXT: CallExpr{{.*}} 'int'
+ // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int (*)()' <FunctionToPointerDecay>
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'int ()' lvalue Function{{.*}} 'returns_3' 'int ()'
+ // CHECK-NEXT: IntegerLiteral{{.*}} 2
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+
+//#pragma acc parallel private(CEArray[: I])
+ // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel
+ // CHECK-NEXT: private clause
+ // CHECK-NEXT: ArraySectionExpr
+ // CHECK-NEXT: ImplicitCastExpr{{.*}}'const int *' <ArrayToPointerDecay>
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'const int[5]' lvalue
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'auto &' depth 0 index 2 CEArray
+ // CHECK-NEXT: DeclRefExpr{{.*}} 'const int[5]' lvalue Var{{.*}}'CEArray' 'const int[5]'
+ // CHECK-NEXT: <<<NULL>>>
+ // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}} 'unsigned int'
+ // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} 'unsigned int' depth 0 index 1 I
+ // CHECK-NEXT: IntegerLiteral{{.*}} 'unsigned int' 3
+ // CHECK-NEXT: WhileStmt
+ // CHECK-NEXT: CXXBoolLiteralExpr
+ // CHECK-NEXT: NullStmt
+}
+
+// CHECK-NEXT: FunctionDecl{{.*}}inst
+void inst() {
+ static constexpr int CEArray[5]={1,2,3,4,5};
+ Templ<int, 3, CEArray>(5);
+}
+#endif
diff --git a/clang/test/SemaOpenACC/sub-array.cpp b/clang/test/SemaOpenACC/sub-array.cpp
new file mode 100644
index 00000000000000..355ac5ef1d3cee
--- /dev/null
+++ b/clang/test/SemaOpenACC/sub-array.cpp
@@ -0,0 +1,208 @@
+// RUN: %clang_cc1 %s -verify -fopenacc
+
+struct Incomplete; // #INCOMPLETE
+struct NotConvertible{} NC;
+
+struct CorrectConvert {
+ operator int();
+} Convert;
+
+constexpr int returns_3() { return 3; }
+
+using FuncPtrTy = void (*)();
+FuncPtrTy FuncPtrTyArray[2];
+
+void Func(int i, int j) {
+ int array[5];
+ int VLA[i];
+ int *ptr;
+ void *void_ptr;
+
+ // Follows int-expr rules, so only convertible to int.
+ // expected-error at +1{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(array[NC:])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(array[:NC])
+ while (true);
+
+ // expected-error at +2{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+ // expected-error at +1{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(array[NC:NC])
+ while (true);
+
+ // expected-error at +2{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel private(ptr[NC:])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(ptr[:NC])
+ while (true);
+
+ // expected-error at +2{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+ // expected-error at +1{{OpenACC sub-array bound requires expression of integer type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(ptr[NC:NC])
+ while (true);
+
+ // These are convertible, so they work.
+#pragma acc parallel private(array[Convert:Convert])
+ while (true);
+
+#pragma acc parallel private(ptr[Convert:Convert])
+ while (true);
+
+
+ // The length for "dynamically" allocated dimensions of an array must be
+ // explicitly specified.
+
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel private(ptr[3:])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is an array of unknown bound}}
+#pragma acc parallel private(VLA[3:])
+ while (true);
+
+#pragma acc parallel private(ptr[:3])
+ while (true);
+
+#pragma acc parallel private(VLA[:3])
+ while (true);
+
+ // Error if the length of the array + the initializer is bigger the the array
+ // with known bounds.
+
+ // expected-error at +1{{OpenACC sub-array length evaluated to a value (6) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(array[i:returns_3() + 3])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array length evaluated to a value (6) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(array[:returns_3() + 3])
+ while (true);
+
+#pragma acc parallel private(array[:returns_3()])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array specified range [3:3] would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(array[returns_3():returns_3()])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array lower bound evaluated to a value (6) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(array[returns_3() + 3:])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array lower bound evaluated to a value (6) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(array[returns_3() + 3:1])
+ while (true);
+
+ // Standard doesn't specify this, but negative values are likely not
+ // permitted, so disallow them here until we come up with a good reason to do
+ // otherwise.
+
+ // expected-error at +1{{OpenACC sub-array lower bound evaluated to negative value -1}}
+#pragma acc parallel private(array[returns_3() - 4 : ])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array length evaluated to negative value -1}}
+#pragma acc parallel private(array[: -1])
+ while (true);
+
+ Incomplete *IncompletePtr;
+ // expected-error at +2{{OpenACC sub-array base is of incomplete type 'Incomplete'}}
+ // expected-note@#INCOMPLETE{{forward declaration of 'Incomplete'}}
+#pragma acc parallel private(IncompletePtr[0 :1])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array base is of incomplete type 'void'}}
+#pragma acc parallel private(void_ptr[0:1])
+ while (true);
+
+ // OK: these are function pointers.
+#pragma acc parallel private(FuncPtrTyArray[0 :1])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array cannot be of function type 'void ()'}}
+#pragma acc parallel private(FuncPtrTyArray[0][0 :1])
+ while (true);
+
+
+ // expected-error at +1{{OpenACC sub-array subscripted value is not an array or pointer}}
+#pragma acc parallel private(i[0:1])
+ while (true);
+}
+
+template<typename T, typename U, typename V, unsigned I, auto &CEArray>
+void Templ(int i){
+ T array[I];
+ T VLA[i];
+ T *ptr;
+ U NC;
+ V Conv;
+
+ // Convertible:
+ // expected-error at +2{{OpenACC sub-array bound requires expression of integer type ('NotConvertible' invalid}}
+ // expected-note@#INST{{in instantiation of function template specialization}}
+#pragma acc parallel private(array[NC:])
+ while (true);
+ // expected-error at +1{{OpenACC sub-array bound requires expression of integer type ('NotConvertible' invalid}}
+#pragma acc parallel private(array[:NC])
+ while (true);
+
+#pragma acc parallel private(array[Conv:])
+ while (true);
+#pragma acc parallel private(array[:Conv])
+ while (true);
+
+ // Need a length for unknown size.
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is not an array}}
+#pragma acc parallel private(ptr[Conv:])
+ while (true);
+ // expected-error at +1{{OpenACC sub-array length is unspecified and cannot be inferred because the subscripted value is an array of unknown bound}}
+#pragma acc parallel private(VLA[Conv:])
+ while (true);
+#pragma acc parallel private(ptr[:Conv])
+ while (true);
+#pragma acc parallel private(VLA[:Conv])
+ while (true);
+
+ // Out of bounds.
+ // expected-error at +1{{OpenACC sub-array lower bound evaluated to a value (2) that would be out of the range of the subscripted array size of 2}}
+#pragma acc parallel private(array[I:])
+ while (true);
+
+ // OK, don't know the value.
+#pragma acc parallel private(array[i:])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array length evaluated to a value (3) that would be out of the range of the subscripted array size of 2}}
+#pragma acc parallel private(array[:I + 1])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array lower bound evaluated to a value (5) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(CEArray[5:])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array length evaluated to a value (6) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(CEArray[:2 + I + I])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array length evaluated to a value (4294967295) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(CEArray[:1 - I])
+ while (true);
+
+ // expected-error at +1{{OpenACC sub-array lower bound evaluated to a value (4294967295) that would be out of the range of the subscripted array size of 5}}
+#pragma acc parallel private(CEArray[1 - I:])
+ while (true);
+
+ T not_ptr;
+ // expected-error at +1{{OpenACC sub-array subscripted value is not an array or pointer}}
+#pragma acc parallel private(not_ptr[0:1])
+ while (true);
+}
+
+void inst() {
+ static constexpr int CEArray[5]={1,2,3,4,5};
+ Templ<int, NotConvertible, CorrectConvert, 2, CEArray>(5); // #INST
+}
>From 3d13ef54ef724cdc6992bb42b8c7ec2d482e50ad Mon Sep 17 00:00:00 2001
From: erichkeane <ekeane at nvidia.com>
Date: Thu, 2 May 2024 06:22:50 -0700
Subject: [PATCH 2/2] Change variable name
---
clang/lib/Sema/SemaOpenACC.cpp | 4 ++--
1 file changed, 2 insertions(+), 2 deletions(-)
diff --git a/clang/lib/Sema/SemaOpenACC.cpp b/clang/lib/Sema/SemaOpenACC.cpp
index a203ece7ff3130..e07c5b2317f279 100644
--- a/clang/lib/Sema/SemaOpenACC.cpp
+++ b/clang/lib/Sema/SemaOpenACC.cpp
@@ -685,8 +685,8 @@ ExprResult SemaOpenACC::ActOnArraySectionExpr(Expr *Base, SourceLocation LBLoc,
if (LHS.isSigned() == RHS.isSigned())
return LHS + RHS;
- unsigned width = std::max(LHS.getBitWidth(), RHS.getBitWidth()) + 1;
- return llvm::APSInt(LHS.sext(width) + RHS.sext(width), /*Signed=*/true);
+ unsigned Width = std::max(LHS.getBitWidth(), RHS.getBitWidth()) + 1;
+ return llvm::APSInt(LHS.sext(Width) + RHS.sext(Width), /*Signed=*/true);
};
// If we know all 3 values, we can diagnose that the total value would be out
More information about the cfe-commits
mailing list