[clang] e777e7a - [OpenACC] Implement better dupe catching for device_type clauses (#138196)
via cfe-commits
cfe-commits at lists.llvm.org
Fri May 2 09:01:34 PDT 2025
Author: Erich Keane
Date: 2025-05-02T09:01:31-07:00
New Revision: e777e7a5d21b2fb3e77f16b86c4531f35a0a5cac
URL: https://github.com/llvm/llvm-project/commit/e777e7a5d21b2fb3e77f16b86c4531f35a0a5cac
DIFF: https://github.com/llvm/llvm-project/commit/e777e7a5d21b2fb3e77f16b86c4531f35a0a5cac.diff
LOG: [OpenACC] Implement better dupe catching for device_type clauses (#138196)
Previously we just checked duplicates for a handful of clauses between
active device_type clauses. However, after looking into the
implementation for 'collapse', I discovered that duplicate device_type
identifiers with duplicate versions of htese clause are problematic.
This patch corrects this and now catches these conflicts.
Added:
Modified:
clang/include/clang/Basic/DiagnosticSemaKinds.td
clang/lib/Sema/SemaOpenACCClause.cpp
clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c
clang/test/SemaOpenACC/compute-construct-num_workers-clause.c
clang/test/SemaOpenACC/compute-construct-vector_length-clause.c
clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
clang/test/SemaOpenACC/loop-construct-tile-clause.cpp
Removed:
################################################################################
diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td
index c630f7851c055..ccb14e9927adf 100644
--- a/clang/include/clang/Basic/DiagnosticSemaKinds.td
+++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td
@@ -13000,6 +13000,13 @@ def err_acc_duplicate_clause_disallowed
: Error<"OpenACC '%1' clause cannot appear more than once on a '%0' "
"directive">;
def note_acc_previous_clause_here : Note<"previous clause is here">;
+// TODO(OpenACC): Combine these with the one above, and decide between a
+// 'select' to split between showing the clause name, or just always printing
+// the name.
+def note_acc_previous_named_clause_here : Note<"previous '%0' clause is here">;
+def note_acc_device_type_here
+ : Note<"%enum_select<ACCDeviceTypeApp>{%Active{active}|%Applies{which "
+ "applies to}}0 'device_type' clause here">;
def note_acc_previous_expr_here : Note<"previous expression is here">;
def note_acc_previous_reference : Note<"previous reference is here">;
def err_acc_branch_in_out_compute_construct
@@ -13085,6 +13092,9 @@ def err_acc_clause_routine_one_of_in_region
def err_acc_clause_since_last_device_type
: Error<"OpenACC '%0' clause cannot appear more than once%select{| in a "
"'device_type' region}2 on a '%1' directive">;
+def err_acc_clause_conflicts_prev_dev_type
+ : Error<"OpenACC '%0' clause applies to 'device_type' '%1', which "
+ "conflicts with previous '%0' clause">;
def err_acc_reduction_num_gangs_conflict
: Error<"OpenACC '%1' clause %select{|with more than 1 argument }0may not "
diff --git a/clang/lib/Sema/SemaOpenACCClause.cpp b/clang/lib/Sema/SemaOpenACCClause.cpp
index 6cf6888e2a3a9..ee0d217de74ff 100644
--- a/clang/lib/Sema/SemaOpenACCClause.cpp
+++ b/clang/lib/Sema/SemaOpenACCClause.cpp
@@ -335,29 +335,106 @@ class SemaOpenACCClauseVisitor {
// For 'tile' and 'collapse', only allow 1 per 'device_type'.
// Also applies to num_worker, num_gangs, vector_length, and async.
+ // This does introspection into the actual device-types to prevent duplicates
+ // across device types as well.
template <typename TheClauseTy>
bool DisallowSinceLastDeviceType(SemaOpenACC::OpenACCParsedClause &Clause) {
auto LastDeviceTypeItr =
std::find_if(ExistingClauses.rbegin(), ExistingClauses.rend(),
llvm::IsaPred<OpenACCDeviceTypeClause>);
- auto Last = std::find_if(ExistingClauses.rbegin(), LastDeviceTypeItr,
- llvm::IsaPred<TheClauseTy>);
+ auto LastSinceDevTy =
+ std::find_if(ExistingClauses.rbegin(), LastDeviceTypeItr,
+ llvm::IsaPred<TheClauseTy>);
- if (Last == LastDeviceTypeItr)
+ // In this case there is a duplicate since the last device_type/lack of a
+ // device_type. Diagnose these as duplicates.
+ if (LastSinceDevTy != LastDeviceTypeItr) {
+ SemaRef.Diag(Clause.getBeginLoc(),
+ diag::err_acc_clause_since_last_device_type)
+ << Clause.getClauseKind() << Clause.getDirectiveKind()
+ << (LastDeviceTypeItr != ExistingClauses.rend());
+ SemaRef.Diag((*LastSinceDevTy)->getBeginLoc(),
+ diag::note_acc_previous_clause_here);
+
+ // Mention the last device_type as well.
+ if (LastDeviceTypeItr != ExistingClauses.rend())
+ SemaRef.Diag((*LastDeviceTypeItr)->getBeginLoc(),
+ diag::note_acc_previous_clause_here);
+ return true;
+ }
+
+ // If this isn't in a device_type, and we didn't diagnose that there are
+ // dupes above, just give up, no sense in searching for previous device_type
+ // regions as they don't exist.
+ if (LastDeviceTypeItr == ExistingClauses.rend())
return false;
- SemaRef.Diag(Clause.getBeginLoc(),
- diag::err_acc_clause_since_last_device_type)
- << Clause.getClauseKind() << Clause.getDirectiveKind()
- << (LastDeviceTypeItr != ExistingClauses.rend());
- SemaRef.Diag((*Last)->getBeginLoc(), diag::note_acc_previous_clause_here);
+ // The device-type that is active for us, so we can compare to the previous
+ // ones.
+ const auto &ActiveDeviceTypeClause =
+ cast<OpenACCDeviceTypeClause>(**LastDeviceTypeItr);
- if (LastDeviceTypeItr != ExistingClauses.rend())
- SemaRef.Diag((*LastDeviceTypeItr)->getBeginLoc(),
- diag::note_acc_previous_clause_here);
+ auto PrevDeviceTypeItr = LastDeviceTypeItr;
+ auto CurDevTypeItr = LastDeviceTypeItr;
- return true;
+ while ((CurDevTypeItr = std::find_if(
+ std::next(PrevDeviceTypeItr), ExistingClauses.rend(),
+ llvm::IsaPred<OpenACCDeviceTypeClause>)) !=
+ ExistingClauses.rend()) {
+ // At this point, we know that we have a region between two device_types,
+ // as specified by CurDevTypeItr and PrevDeviceTypeItr.
+
+ auto CurClauseKindItr = std::find_if(PrevDeviceTypeItr, CurDevTypeItr,
+ llvm::IsaPred<TheClauseTy>);
+
+ // There are no clauses of the current kind between these device_types, so
+ // continue.
+ if (CurClauseKindItr == CurDevTypeItr)
+ continue;
+
+ // At this point, we know that this device_type region has a collapse. So
+ // diagnose if the two device_types have any overlap in their
+ // architectures.
+ const auto &CurDeviceTypeClause =
+ cast<OpenACCDeviceTypeClause>(**CurDevTypeItr);
+
+ for (const DeviceTypeArgument &arg :
+ ActiveDeviceTypeClause.getArchitectures()) {
+ for (const DeviceTypeArgument &prevArg :
+ CurDeviceTypeClause.getArchitectures()) {
+
+ // This should catch duplicates * regions, duplicate same-text (thanks
+ // to identifier equiv.) and case insensitive dupes.
+ if (arg.getIdentifierInfo() == prevArg.getIdentifierInfo() ||
+ (arg.getIdentifierInfo() && prevArg.getIdentifierInfo() &&
+ StringRef{arg.getIdentifierInfo()->getName()}.equals_insensitive(
+ prevArg.getIdentifierInfo()->getName()))) {
+ SemaRef.Diag(Clause.getBeginLoc(),
+ diag::err_acc_clause_conflicts_prev_dev_type)
+ << Clause.getClauseKind()
+ << (arg.getIdentifierInfo() ? arg.getIdentifierInfo()->getName()
+ : "*");
+ // mention the active device type.
+ SemaRef.Diag(ActiveDeviceTypeClause.getBeginLoc(),
+ diag::note_acc_device_type_here)
+ << diag::ACCDeviceTypeApp::Active;
+ // mention the previous clause.
+ SemaRef.Diag((*CurClauseKindItr)->getBeginLoc(),
+ diag::note_acc_previous_named_clause_here)
+ << (*CurClauseKindItr)->getClauseKind();
+ // mention the previous device type.
+ SemaRef.Diag(CurDeviceTypeClause.getBeginLoc(),
+ diag::note_acc_device_type_here)
+ << diag::ACCDeviceTypeApp::Applies;
+ return true;
+ }
+ }
+ }
+
+ PrevDeviceTypeItr = CurDevTypeItr;
+ }
+ return false;
}
public:
diff --git a/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c b/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c
index 9651514ca0fff..904928ce7248b 100644
--- a/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c
@@ -71,3 +71,18 @@ void Test() {
#pragma acc loop num_gangs(1)
for(int i = 5; i < 10;++i);
}
+
+void no_dupes_since_last_device_type() {
+ // expected-error at +4{{OpenACC 'num_gangs' clause applies to 'device_type' 'radeon', which conflicts with previous 'num_gangs' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'num_gangs' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc parallel device_type(nvidia, radeon) num_gangs(getS()) device_type(radeon) num_gangs(getS())
+ ;
+ // expected-error at +4{{OpenACC 'num_gangs' clause applies to 'device_type' 'nvidia', which conflicts with previous 'num_gangs' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'num_gangs' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc parallel device_type(nvidia) num_gangs(getS()) device_type(nvidia, radeon) num_gangs(getS())
+ ;
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c b/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c
index 23a73f0bf49bd..cea0fdc85d9b5 100644
--- a/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-num_workers-clause.c
@@ -60,3 +60,18 @@ void Test() {
#pragma acc loop num_workers(1)
for(int i = 5; i < 10;++i);
}
+
+void no_dupes_since_last_device_type() {
+ // expected-error at +4{{OpenACC 'num_workers' clause applies to 'device_type' 'radEon', which conflicts with previous 'num_workers' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'num_workers' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc parallel device_type(nvidia, radeon) num_workers(getS()) device_type(radEon) num_workers(getS())
+ ;
+ // expected-error at +4{{OpenACC 'num_workers' clause applies to 'device_type' 'nvidia', which conflicts with previous 'num_workers' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'num_workers' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc parallel device_type(nvidia) num_workers(getS()) device_type(nvidia, radeon) num_workers(getS())
+ ;
+}
diff --git a/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c
index c79fc627252f1..5b678593fd3cf 100644
--- a/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c
+++ b/clang/test/SemaOpenACC/compute-construct-vector_length-clause.c
@@ -60,3 +60,18 @@ void Test() {
#pragma acc loop vector_length(1)
for(int i = 5; i < 10;++i);
}
+
+void no_dupes_since_last_device_type() {
+ // expected-error at +4{{OpenACC 'vector_length' clause applies to 'device_type' 'radeon', which conflicts with previous 'vector_length' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'vector_length' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc parallel device_type(nvidia, radeon) vector_length(getS()) device_type(radeon) vector_length(getS())
+ ;
+ // expected-error at +4{{OpenACC 'vector_length' clause applies to 'device_type' 'nvidia', which conflicts with previous 'vector_length' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'vector_length' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc parallel device_type(nvidia) vector_length(getS()) device_type(nvidia, radeon) vector_length(getS())
+ ;
+}
diff --git a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
index 851742434a1f9..cbaf414c04453 100644
--- a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
+++ b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp
@@ -539,4 +539,37 @@ void no_dupes_since_last_device_type() {
#pragma acc loop collapse(1) device_type(*) collapse(1) device_type(nvidia) collapse(2)
for(unsigned i = 0; i < 5; ++i)
for(unsigned j = 0; j < 5; ++j);
+
+ // This one is ok, despite * being the 'all' value.
+#pragma acc loop device_type(*) collapse(1) device_type(nvidia) collapse(2)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+
+#pragma acc loop device_type(nvidia) collapse(1) device_type(*) collapse(2)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+
+ // expected-error at +4{{OpenACC 'collapse' clause applies to 'device_type' '*', which conflicts with previous 'collapse' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'collapse' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc loop device_type(*) collapse(1) device_type(*) collapse(2)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+
+ // expected-error at +4{{OpenACC 'collapse' clause applies to 'device_type' 'nvidia', which conflicts with previous 'collapse' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'collapse' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc loop device_type(nviDia, radeon) collapse(1) device_type(nvidia) collapse(2)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+
+ // expected-error at +4{{OpenACC 'collapse' clause applies to 'device_type' 'radeon', which conflicts with previous 'collapse' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'collapse' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc loop device_type(radeon) collapse(1) device_type(nvidia, radeon) collapse(2)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
}
diff --git a/clang/test/SemaOpenACC/loop-construct-tile-clause.cpp b/clang/test/SemaOpenACC/loop-construct-tile-clause.cpp
index ebd657791f34d..a5b24024d3e5b 100644
--- a/clang/test/SemaOpenACC/loop-construct-tile-clause.cpp
+++ b/clang/test/SemaOpenACC/loop-construct-tile-clause.cpp
@@ -419,4 +419,20 @@ void no_dupes_since_last_device_type() {
#pragma acc loop tile(1) device_type(*) tile(1) device_type(nvidia) tile(2)
for(unsigned i = 0; i < 5; ++i)
for(unsigned j = 0; j < 5; ++j);
+
+ // expected-error at +4{{OpenACC 'tile' clause applies to 'device_type' 'nvidiA', which conflicts with previous 'tile' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'tile' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc loop device_type(nvidia, radeon) tile(1) device_type(nvidiA) tile(2)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
+
+ // expected-error at +4{{OpenACC 'tile' clause applies to 'device_type' 'radeon', which conflicts with previous 'tile' clause}}
+ // expected-note at +3{{active 'device_type' clause here}}
+ // expected-note at +2{{previous 'tile' clause is here}}
+ // expected-note at +1{{which applies to 'device_type' clause here}}
+#pragma acc loop device_type(radeon) tile(1) device_type(nvidia, radeon) tile(2)
+ for(unsigned i = 0; i < 5; ++i)
+ for(unsigned j = 0; j < 5; ++j);
}
More information about the cfe-commits
mailing list