https://github.com/erichkeane updated
https://github.com/llvm/llvm-project/pull/90796
>From ac69522bf15c7c169c4cdb9d3d1547c5c0962193 Mon Sep 17 00:00:00 2001
From: erichkeane <eke...@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@+2{{expected expression}}
// expected-warning@+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@+2{{OpenACC sub-array is not allowed here}}
// expected-warning@+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@+2{{OpenACC sub-array is not allowed here}}
// expected-warning@+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@+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@+2{{expected expression}}
+ // expected-error@+2{{OpenACC sub-array is not allowed here}}
// expected-warning@+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@+2{{expected expression}}
+ // expected-error@+2{{OpenACC sub-array is not allowed here}}
// expected-warning@+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@+3{{expected expression}}
// expected-warning@+2{{OpenACC clause 'copy' not yet implemented, clause
ignored}}
// expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause
ignored}}
#pragma acc serial copy(HasMem.MemArr[:]), seq
@@ -519,7 +518,6 @@ void VarListClauses() {
#pragma acc serial copy(HasMem.MemArr[: :]), seq
for(;;){}
- // expected-error@+3{{expected expression}}
// expected-warning@+2{{OpenACC clause 'copy' not yet implemented, clause
ignored}}
// expected-warning@+1{{OpenACC clause 'seq' not yet implemented, clause
ignored}}
#pragma acc serial copy(HasMem.MemArr[3:]), seq
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@+1{{expected unqualified-id}}
+#pragma acc parallel private(array[::])
+ while (true);
+ // expected-error@+2{{expected ']'}}
+ // expected-note@+1{{to match this '['}}
+#pragma acc parallel private(array[1::])
+ while (true);
+ // expected-error@+2{{expected ']'}}
+ // expected-note@+1{{to match this '['}}
+#pragma acc parallel private(array[:2:])
+ while (true);
+ // expected-error@+3{{expected unqualified-id}}
+ // expected-error@+2{{expected ']'}}
+ // expected-note@+1{{to match this '['}}
+#pragma acc parallel private(array[::3])
+ while (true);
+ // expected-error@+2{{expected ']'}}
+ // expected-note@+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@+1{{expected unqualified-id}}
+#pragma acc parallel private(array[::])
+ while (true);
+ // expected-error@+3{{'t' is not a class, namespace, or enumeration}}
+ // expected-note@#tDecl{{'t' declared here}}
+ // expected-error@+1{{expected unqualified-id}}
+#pragma acc parallel private(array[t::])
+ while (true);
+ // expected-error@+2{{expected ']'}}
+ // expected-note@+1{{to match this '['}}
+#pragma acc parallel private(array[:I:])
+ while (true);
+ // expected-error@+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@+2{{expected ']'}}
+ // expected-note@+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@+1{{expected expression}}
#pragma acc parallel private(LocalArray[:])
while(1);
#pragma acc parallel private(LocalArray[:5])
while(1);
- // TODO OpenACC: Fix array sections, this should be allowed.
- // expected-error@+1{{expected expression}}
#pragma acc parallel private(LocalArray[2:])
while(1);
-#pragma acc parallel private(LocalArray[2:5])
+#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@+1{{expected expression}}
+ // expected-error@+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@+1{{expected expression}}
+ // expected-error@+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@+1{{expected expression}}
#pragma acc parallel private(ArrayParam[:])
while(1);
#pragma acc parallel private(ArrayParam[:5])
while(1);
#pragma acc parallel private(ArrayParam[:IntParam])
while(1);
- // TODO OpenACC: Fix array sections, this should be allowed.
- // expected-error@+1{{expected expression}}
#pragma acc parallel private(ArrayParam[2:])
while(1);
+ // expected-error@+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@+1{{OpenACC sub-array is not allowed here}}
+ // expected-error@+2{{OpenACC sub-array specified range [2:5] would be out
of the range of the subscripted array size of 5}}
+ // expected-error@+1{{OpenACC variable is not a valid variable name,
sub-array, array element, or composite variable member}}
#pragma acc parallel private((float*)ArrayParam[2:5])
while(1);
// expected-error@+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@+1{{expected expression}}
+ // expected-error@+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@+1{{OpenACC sub-array bound requires expression of integer
type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(array[NC:])
+ while (true);
+
+ // expected-error@+1{{OpenACC sub-array bound requires expression of integer
type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(array[:NC])
+ while (true);
+
+ // expected-error@+2{{OpenACC sub-array bound requires expression of integer
type ('struct NotConvertible' invalid}}
+ // expected-error@+1{{OpenACC sub-array bound requires expression of integer
type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(array[NC:NC])
+ while (true);
+
+ // expected-error@+2{{OpenACC sub-array bound requires expression of integer
type ('struct NotConvertible' invalid}}
+ // expected-error@+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@+1{{OpenACC sub-array bound requires expression of integer
type ('struct NotConvertible' invalid}}
+#pragma acc parallel private(ptr[:NC])
+ while (true);
+
+ // expected-error@+2{{OpenACC sub-array bound requires expression of integer
type ('struct NotConvertible' invalid}}
+ // expected-error@+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@+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@+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@+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@+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@+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@+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@+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@+1{{OpenACC sub-array lower bound evaluated to negative
value -1}}
+#pragma acc parallel private(array[returns_3() - 4 : ])
+ while (true);
+
+ // expected-error@+1{{OpenACC sub-array length evaluated to negative value
-1}}
+#pragma acc parallel private(array[: -1])
+ while (true);
+
+ Incomplete *IncompletePtr;
+ // expected-error@+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@+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@+1{{OpenACC sub-array cannot be of function type 'void
()'}}
+#pragma acc parallel private(FuncPtrTyArray[0][0 :1])
+ while (true);
+
+
+ // expected-error@+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@+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@+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@+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@+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@+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@+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@+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@+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@+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@+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@+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 <eke...@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
_______________________________________________
cfe-commits mailing list
cfe-commits@lists.llvm.org
https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
