[clang] [OpenACC} Implement SubArray Parsing/Sema (PR #90796)
via cfe-commits
cfe-commits at lists.llvm.org
Wed May 1 16:04:26 PDT 2024
llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT-->
@llvm/pr-subscribers-clang
Author: Erich Keane (erichkeane)
<details>
<summary>Changes</summary>
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.
---
Patch is 60.81 KiB, truncated to 20.00 KiB below, full version: https://github.com/llvm/llvm-project/pull/90796.diff
12 Files Affected:
- (modified) clang/include/clang/Basic/DiagnosticSemaKinds.td (+22-2)
- (modified) clang/lib/Parse/ParseExpr.cpp (+2-1)
- (modified) clang/lib/Sema/SemaOpenACC.cpp (+220-12)
- (modified) clang/test/ParserOpenACC/parse-cache-construct.c (+1-2)
- (modified) clang/test/ParserOpenACC/parse-cache-construct.cpp (+4-4)
- (modified) clang/test/ParserOpenACC/parse-clauses.c (-2)
- (added) clang/test/ParserOpenACC/parse-sub-array.cpp (+89)
- (modified) clang/test/SemaOpenACC/compute-construct-private-clause.c (+9-17)
- (modified) clang/test/SemaOpenACC/compute-construct-private-clause.cpp (+1-2)
- (modified) clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp (+9)
- (added) clang/test/SemaOpenACC/sub-array-ast.cpp (+566)
- (added) clang/test/SemaOpenACC/sub-array.cpp (+208)
``````````diff
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' declare...
[truncated]
``````````
</details>
https://github.com/llvm/llvm-project/pull/90796
More information about the cfe-commits
mailing list