Author: erichkeane Date: 2025-06-26T10:26:20-07:00 New Revision: 0a2b6f6c1cc796bcac38d1c6b44d76de6049ea12
URL: https://github.com/llvm/llvm-project/commit/0a2b6f6c1cc796bcac38d1c6b44d76de6049ea12 DIFF: https://github.com/llvm/llvm-project/commit/0a2b6f6c1cc796bcac38d1c6b44d76de6049ea12.diff LOG: [OpenACC] Fix 'copyout' allowed modifiers alwaysin vs alwaysout While doing lowering, I discovered that the restriction onthe allowed modifiers for 'copyout' didn't make sense! After discussion on the OpenACC standards mailing list I discovered that this was a copy/paste error during standardization that they intend to fix, and really meant for copyout to allow alwaysout instead of alwaysin. When implementing, I blindly followed the standard :) This patch corrects the implementation to do what was meant. Added: Modified: clang/lib/Sema/SemaOpenACCClause.cpp clang/test/AST/ast-print-openacc-combined-construct.cpp clang/test/SemaOpenACC/combined-construct-copyout-ast.cpp clang/test/SemaOpenACC/combined-construct-copyout-clause.c clang/test/SemaOpenACC/compute-construct-copyout-clause.c clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp clang/test/SemaOpenACC/data-construct-copyout-ast.cpp clang/test/SemaOpenACC/data-construct-copyout-clause.c clang/test/SemaOpenACC/declare-construct.cpp Removed: ################################################################################ diff --git a/clang/lib/Sema/SemaOpenACCClause.cpp b/clang/lib/Sema/SemaOpenACCClause.cpp index 8ee27694b7376..3f90fe8ff8438 100644 --- a/clang/lib/Sema/SemaOpenACCClause.cpp +++ b/clang/lib/Sema/SemaOpenACCClause.cpp @@ -231,8 +231,8 @@ class SemaOpenACCClauseVisitor { case OpenACCClauseKind::PCopyOut: case OpenACCClauseKind::PresentOrCopyOut: // COPYOUT: Capture only struct.data & compute - return Check(OpenACCModifierKind::Always | OpenACCModifierKind::AlwaysIn | - OpenACCModifierKind::Zero | + return Check(OpenACCModifierKind::Always | + OpenACCModifierKind::AlwaysOut | OpenACCModifierKind::Zero | (IsStructuredDataOrCompute ? OpenACCModifierKind::Capture : OpenACCModifierKind::Invalid)); case OpenACCClauseKind::Create: diff --git a/clang/test/AST/ast-print-openacc-combined-construct.cpp b/clang/test/AST/ast-print-openacc-combined-construct.cpp index 7331ec3786e99..b4e803348a2f7 100644 --- a/clang/test/AST/ast-print-openacc-combined-construct.cpp +++ b/clang/test/AST/ast-print-openacc-combined-construct.cpp @@ -179,8 +179,8 @@ void foo() { #pragma acc parallel loop copyin(i, array[1], array, array[1:2]) pcopyin(readonly:i, array[1], array, array[1:2]) present_or_copyin(always, alwaysin: i, array[1], array, array[1:2]) for(int i = 0;i<5;++i); -// CHECK: #pragma acc parallel loop copyout(i, array[1], array, array[1:2]) pcopyout(zero: i, array[1], array, array[1:2]) present_or_copyout(always, alwaysin: i, array[1], array, array[1:2]) -#pragma acc parallel loop copyout(i, array[1], array, array[1:2]) pcopyout(zero: i, array[1], array, array[1:2]) present_or_copyout(always, alwaysin: i, array[1], array, array[1:2]) +// CHECK: #pragma acc parallel loop copyout(i, array[1], array, array[1:2]) pcopyout(zero: i, array[1], array, array[1:2]) present_or_copyout(always, alwaysout: i, array[1], array, array[1:2]) +#pragma acc parallel loop copyout(i, array[1], array, array[1:2]) pcopyout(zero: i, array[1], array, array[1:2]) present_or_copyout(always, alwaysout: i, array[1], array, array[1:2]) for(int i = 0;i<5;++i); // CHECK: #pragma acc parallel loop create(i, array[1], array, array[1:2]) pcreate(zero: i, array[1], array, array[1:2]) present_or_create(i, array[1], array, array[1:2]) diff --git a/clang/test/SemaOpenACC/combined-construct-copyout-ast.cpp b/clang/test/SemaOpenACC/combined-construct-copyout-ast.cpp index ec4451e9df7d3..f865f826757eb 100644 --- a/clang/test/SemaOpenACC/combined-construct-copyout-ast.cpp +++ b/clang/test/SemaOpenACC/combined-construct-copyout-ast.cpp @@ -14,7 +14,7 @@ void NormalUses(float *PointerParam) { // CHECK: ParmVarDecl // CHECK-NEXT: CompoundStmt -#pragma acc parallel loop copyout(GlobalArray) pcopyout(zero:PointerParam[Global]) present_or_copyout(always, alwaysin: Global) +#pragma acc parallel loop copyout(GlobalArray) pcopyout(zero:PointerParam[Global]) present_or_copyout(always, alwaysout: Global) for (unsigned i = 0; i < 5; ++i); // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop // CHECK-NEXT: copyout clause @@ -25,7 +25,7 @@ void NormalUses(float *PointerParam) { // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *' // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue> // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' - // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysin + // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysout // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' // CHECK-NEXT: For // CHECK: NullStmt @@ -42,10 +42,10 @@ void TemplUses(T t, U u) { // CHECK-NEXT: ParmVarDecl{{.*}} referenced u 'U' // CHECK-NEXT: CompoundStmt -#pragma acc parallel loop copyout(always, alwaysin: t) pcopyout(zero: NTTP, u) present_or_copyout(u[0:t]) +#pragma acc parallel loop copyout(always, alwaysout: t) pcopyout(zero: NTTP, u) present_or_copyout(u[0:t]) for (unsigned i = 0; i < 5; ++i); // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop - // CHECK-NEXT: copyout clause modifiers: always, alwaysin + // CHECK-NEXT: copyout clause modifiers: always, alwaysout // CHECK-NEXT: DeclRefExpr{{.*}}'T' lvalue ParmVar{{.*}} 't' 'T' // CHECK-NEXT: pcopyout clause modifiers: zero // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &' @@ -73,7 +73,7 @@ void TemplUses(T t, U u) { // #pragma acc parallel copyout(t) pcopyout(zero: NTTP, u) present_or_copyout(u[0:t]) // CHECK-NEXT: OpenACCCombinedConstruct{{.*}} parallel loop - // CHECK-NEXT: copyout clause modifiers: always, alwaysin + // CHECK-NEXT: copyout clause modifiers: always, alwaysout // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int' // CHECK-NEXT: pcopyout clause modifiers: zero // CHECK-NEXT: SubstNonTypeTemplateParmExpr{{.*}}'const unsigned int' lvalue diff --git a/clang/test/SemaOpenACC/combined-construct-copyout-clause.c b/clang/test/SemaOpenACC/combined-construct-copyout-clause.c index 2ca70736d30e3..a1774e070a7b0 100644 --- a/clang/test/SemaOpenACC/combined-construct-copyout-clause.c +++ b/clang/test/SemaOpenACC/combined-construct-copyout-clause.c @@ -78,19 +78,19 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo } void ModList() { int V1; - // expected-error@+2{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} + // expected-error@+2{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc parallel loop copyout(always, alwaysin, alwaysout, zero, readonly: V1) for(int i = 0; i < 6;++i); - // expected-error@+1{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} -#pragma acc serial loop copyout(alwaysout: V1) + // expected-error@+1{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} +#pragma acc serial loop copyout(alwaysin: V1) for(int i = 0; i < 6;++i); // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc kernels loop copyout(readonly: V1) for(int i = 0; i < 6;++i); #pragma acc parallel loop copyout(capture:V1) for(int i = 5; i < 10;++i); -#pragma acc parallel loop copyout(always, alwaysin, zero, capture: V1) +#pragma acc parallel loop copyout(always, alwaysout, zero, capture: V1) for(int i = 5; i < 10;++i); } diff --git a/clang/test/SemaOpenACC/compute-construct-copyout-clause.c b/clang/test/SemaOpenACC/compute-construct-copyout-clause.c index 34f8d5aa60e09..c0ab6c3ea6587 100644 --- a/clang/test/SemaOpenACC/compute-construct-copyout-clause.c +++ b/clang/test/SemaOpenACC/compute-construct-copyout-clause.c @@ -78,18 +78,18 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo } void ModList() { int V1; - // expected-error@+2{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} + // expected-error@+2{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc parallel copyout(always, alwaysin, alwaysout, zero, readonly: V1) for(int i = 0; i < 6;++i); - // expected-error@+1{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} -#pragma acc serial copyout(alwaysout: V1) + // expected-error@+1{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} +#pragma acc serial copyout(alwaysin: V1) for(int i = 0; i < 6;++i); // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc kernels copyout(readonly: V1) for(int i = 0; i < 6;++i); #pragma acc parallel copyout(capture:V1) for(int i = 5; i < 10;++i); -#pragma acc parallel copyout(always, alwaysin, zero, capture: V1) +#pragma acc parallel copyout(always, alwaysout, zero, capture: V1) for(int i = 5; i < 10;++i); } diff --git a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp index df01c59349665..e3c41849eba81 100644 --- a/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp +++ b/clang/test/SemaOpenACC/compute-construct-varlist-ast.cpp @@ -133,7 +133,7 @@ void NormalUses(float *PointerParam) { // CHECK-NEXT: CXXBoolLiteralExpr // CHECK-NEXT: NullStmt -#pragma acc parallel copyout(GlobalArray) pcopyout(zero:PointerParam[Global]) present_or_copyout(always, alwaysin: Global) +#pragma acc parallel copyout(GlobalArray) pcopyout(zero:PointerParam[Global]) present_or_copyout(always, alwaysout: Global) while(true); // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel // CHECK-NEXT: copyout clause @@ -144,7 +144,7 @@ void NormalUses(float *PointerParam) { // CHECK-NEXT: DeclRefExpr{{.*}}'float *' lvalue ParmVar{{.*}}'PointerParam' 'float *' // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int' <LValueToRValue> // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' - // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysin + // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysout // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue Var{{.*}}'Global' 'int' // CHECK-NEXT: WhileStmt // CHECK-NEXT: CXXBoolLiteralExpr @@ -363,7 +363,7 @@ void TemplUses(T t, U u, T*PointerParam) { // CHECK-NEXT: CXXBoolLiteralExpr // CHECK-NEXT: NullStmt -#pragma acc parallel copyout(t) pcopyout(zero:NTTP, u) present_or_copyout(always, alwaysin: u[0:t]) +#pragma acc parallel copyout(t) pcopyout(zero:NTTP, u) present_or_copyout(always, alwaysout: u[0:t]) while(true); // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel // CHECK-NEXT: copyout clause @@ -371,7 +371,7 @@ void TemplUses(T t, U u, T*PointerParam) { // CHECK-NEXT: pcopyout clause modifiers: zero // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &' // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' - // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysin + // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysout // CHECK-NEXT: ArraySectionExpr // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0 @@ -569,7 +569,7 @@ void TemplUses(T t, U u, T*PointerParam) { // CHECK-NEXT: CXXBoolLiteralExpr // CHECK-NEXT: NullStmt -// #pragma acc parallel copyout(t) pcopyout(zero:NTTP, u) present_or_copyout(always, alwaysin: u[0:t]) +// #pragma acc parallel copyout(t) pcopyout(zero:NTTP, u) present_or_copyout(always, alwaysout: u[0:t]) // CHECK-NEXT: OpenACCComputeConstruct{{.*}} parallel // CHECK-NEXT: copyout clause // CHECK-NEXT: DeclRefExpr{{.*}}'int' lvalue ParmVar{{.*}} 't' 'int' @@ -578,7 +578,7 @@ void TemplUses(T t, U u, T*PointerParam) { // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int' // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' - // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysin + // CHECK-NEXT: present_or_copyout clause modifiers: always, alwaysout // CHECK-NEXT: ArraySectionExpr // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue> // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' diff --git a/clang/test/SemaOpenACC/data-construct-copyout-ast.cpp b/clang/test/SemaOpenACC/data-construct-copyout-ast.cpp index 069ced7de83d8..1be6854ac996f 100644 --- a/clang/test/SemaOpenACC/data-construct-copyout-ast.cpp +++ b/clang/test/SemaOpenACC/data-construct-copyout-ast.cpp @@ -54,7 +54,7 @@ void TemplUses(T t, U u) { // CHECK-NEXT: ParmVarDecl{{.*}} referenced u 'U' // CHECK-NEXT: CompoundStmt -#pragma acc data copyout(t) pcopyout(zero: NTTP, u) present_or_copyout(alwaysin: u[0:t]) +#pragma acc data copyout(t) pcopyout(zero: NTTP, u) present_or_copyout(alwaysout: u[0:t]) ; // CHECK-NEXT: OpenACCDataConstruct{{.*}} data // CHECK-NEXT: copyout clause @@ -62,7 +62,7 @@ void TemplUses(T t, U u) { // CHECK-NEXT: pcopyout clause modifiers: zero // CHECK-NEXT: DeclRefExpr{{.*}}'auto' lvalue NonTypeTemplateParm{{.*}} 'NTTP' 'auto &' // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' - // CHECK-NEXT: present_or_copyout clause modifiers: alwaysin + // CHECK-NEXT: present_or_copyout clause modifiers: alwaysout // CHECK-NEXT: ArraySectionExpr // CHECK-NEXT: DeclRefExpr{{.*}}'U' lvalue ParmVar{{.*}} 'u' 'U' // CHECK-NEXT: IntegerLiteral{{.*}} 'int' 0 @@ -103,7 +103,7 @@ void TemplUses(T t, U u) { // CHECK-NEXT: NonTypeTemplateParmDecl{{.*}} referenced 'auto &' depth 0 index 0 NTTP // CHECK-NEXT: DeclRefExpr{{.*}}'const unsigned int' lvalue Var{{.*}} 'CEVar' 'const unsigned int' // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' - // CHECK-NEXT: present_or_copyout clause modifiers: alwaysin + // CHECK-NEXT: present_or_copyout clause modifiers: alwaysout // CHECK-NEXT: ArraySectionExpr // CHECK-NEXT: ImplicitCastExpr{{.*}} 'int *' <LValueToRValue> // CHECK-NEXT: DeclRefExpr{{.*}}'int *' lvalue ParmVar{{.*}} 'u' 'int *' diff --git a/clang/test/SemaOpenACC/data-construct-copyout-clause.c b/clang/test/SemaOpenACC/data-construct-copyout-clause.c index deaf48fc77597..79a2f7f4d2634 100644 --- a/clang/test/SemaOpenACC/data-construct-copyout-clause.c +++ b/clang/test/SemaOpenACC/data-construct-copyout-clause.c @@ -74,20 +74,22 @@ void uses(int IntParam, short *PointerParam, float ArrayParam[5], Complete Compo void ModList() { int V1; - // expected-error@+2{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} + // expected-error@+2{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc data copyout(always, alwaysin, alwaysout, zero, readonly: V1) - // expected-error@+1{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} + // expected-error@+1{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} +#pragma acc data copyout(alwaysin: V1) #pragma acc data copyout(alwaysout: V1) // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc data copyout(readonly: V1) #pragma acc data copyout(capture: V1) -#pragma acc data copyout(always, alwaysin, zero, capture: V1) +#pragma acc data copyout(always, alwaysout, zero, capture: V1) - // expected-error@+2{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} + // expected-error@+2{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc exit data copyout(always, alwaysin, alwaysout, zero, readonly: V1) - // expected-error@+1{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} + // expected-error@+1{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} +#pragma acc exit data copyout(alwaysin: V1) #pragma acc exit data copyout(alwaysout: V1) // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc exit data copyout(readonly: V1) diff --git a/clang/test/SemaOpenACC/declare-construct.cpp b/clang/test/SemaOpenACC/declare-construct.cpp index ff8a2ec64c31d..6f21aedc39b45 100644 --- a/clang/test/SemaOpenACC/declare-construct.cpp +++ b/clang/test/SemaOpenACC/declare-construct.cpp @@ -331,17 +331,18 @@ void ModList() { // expected-error@+1{{OpenACC 'capture' modifier not valid on 'copyin' clause}} #pragma acc declare copyin(always, alwaysin, readonly, capture: V8) - // expected-error@+2{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} + // expected-error@+2{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc declare copyout(always, alwaysin, alwaysout, zero, readonly: V9) - // expected-error@+1{{OpenACC 'alwaysout' modifier not valid on 'copyout' clause}} -#pragma acc declare copyout(alwaysout: V10) + // expected-error@+1{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} +#pragma acc declare copyout(alwaysin: V10) // expected-error@+1{{OpenACC 'readonly' modifier not valid on 'copyout' clause}} #pragma acc declare copyout(readonly: V11) // expected-error@+1{{OpenACC 'capture' modifier not valid on 'copyout' clause}} #pragma acc declare copyout(capture: V11B) + // expected-error@+2{{OpenACC 'alwaysin' modifier not valid on 'copyout' clause}} // expected-error@+1{{OpenACC 'capture' modifier not valid on 'copyout' clause}} -#pragma acc declare copyout(always, alwaysin, zero, capture: V12) +#pragma acc declare copyout(always, alwaysin, alwaysout, zero, capture: V12) // expected-error@+5{{OpenACC 'always' modifier not valid on 'create' clause}} // expected-error@+4{{OpenACC 'alwaysin' modifier not valid on 'create' clause}} _______________________________________________ cfe-commits mailing list cfe-commits@lists.llvm.org https://lists.llvm.org/cgi-bin/mailman/listinfo/cfe-commits
