llvmbot wrote:
<!--LLVM PR SUMMARY COMMENT--> @llvm/pr-subscribers-clang Author: Erich Keane (erichkeane) <details> <summary>Changes</summary> 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. --- Full diff: https://github.com/llvm/llvm-project/pull/138196.diff 7 Files Affected: - (modified) clang/include/clang/Basic/DiagnosticSemaKinds.td (+3) - (modified) clang/lib/Sema/SemaOpenACCClause.cpp (+86-12) - (modified) clang/test/SemaOpenACC/compute-construct-num_gangs-clause.c (+15) - (modified) clang/test/SemaOpenACC/compute-construct-num_workers-clause.c (+15) - (modified) clang/test/SemaOpenACC/compute-construct-vector_length-clause.c (+15) - (modified) clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp (+21) - (modified) clang/test/SemaOpenACC/loop-construct-tile-clause.cpp (+16) ``````````diff diff --git a/clang/include/clang/Basic/DiagnosticSemaKinds.td b/clang/include/clang/Basic/DiagnosticSemaKinds.td index 604bb56d28252..f279d84136562 100644 --- a/clang/include/clang/Basic/DiagnosticSemaKinds.td +++ b/clang/include/clang/Basic/DiagnosticSemaKinds.td @@ -13075,6 +13075,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 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..87dae457ac3e9 100644 --- a/clang/lib/Sema/SemaOpenACCClause.cpp +++ b/clang/lib/Sema/SemaOpenACCClause.cpp @@ -335,29 +335,103 @@ 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_previous_clause_here); + // mention the previous clause. + SemaRef.Diag((*CurClauseKindItr)->getBeginLoc(), + diag::note_acc_previous_clause_here); + // mention the previous device type. + SemaRef.Diag(CurDeviceTypeClause.getBeginLoc(), + diag::note_acc_previous_clause_here); + 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..9db6f461081c6 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@+4{{OpenACC 'num_gangs' clause applies to 'device_type' 'radeon', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is here}} +#pragma acc parallel device_type(nvidia, radeon) num_gangs(getS()) device_type(radeon) num_gangs(getS()) + ; + // expected-error@+4{{OpenACC 'num_gangs' clause applies to 'device_type' 'nvidia', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is 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..937d15cb1b65b 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@+4{{OpenACC 'num_workers' clause applies to 'device_type' 'radEon', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is here}} +#pragma acc parallel device_type(nvidia, radeon) num_workers(getS()) device_type(radEon) num_workers(getS()) + ; + // expected-error@+4{{OpenACC 'num_workers' clause applies to 'device_type' 'nvidia', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is 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..e0d7571af687f 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@+4{{OpenACC 'vector_length' clause applies to 'device_type' 'radeon', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is here}} +#pragma acc parallel device_type(nvidia, radeon) vector_length(getS()) device_type(radeon) vector_length(getS()) + ; + // expected-error@+4{{OpenACC 'vector_length' clause applies to 'device_type' 'nvidia', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is 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..ea100d8290a19 100644 --- a/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp +++ b/clang/test/SemaOpenACC/loop-construct-collapse-clause.cpp @@ -539,4 +539,25 @@ 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); + + // expected-error@+4{{OpenACC 'collapse' clause applies to 'device_type' 'nvidia', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is 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@+4{{OpenACC 'collapse' clause applies to 'device_type' 'radeon', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is 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..3fe7e0582b77e 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@+4{{OpenACC 'tile' clause applies to 'device_type' 'nvidiA', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is 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@+4{{OpenACC 'tile' clause applies to 'device_type' 'radeon', which conflicts with previous clause}} + // expected-note@+3{{previous clause is here}} + // expected-note@+2{{previous clause is here}} + // expected-note@+1{{previous clause is 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); } `````````` </details> https://github.com/llvm/llvm-project/pull/138196 _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
