[OpenACC] Fix issue with array-section type generation

[erichkeane]

When an array section bound was in a list of bound-types we improperly generated the list of types of the bounds because
getPointeeOrElementType gets the LOWEST level of type (that is, digs
through ALL array types to get to the base-est of types) when what
we really wanted was 1 layer of pointer/array removed.

This patch fixes it and adds a test that showed the problem by re-ordering the existing ones. This wasn't previously obvious by chance, since the 'array-index-only' variants ended up generating the recipe, and not the bounds.

[llvmbot]

@llvm/pr-subscribers-clangir

@llvm/pr-subscribers-clang

Author: Erich Keane (erichkeane)

Changes

When an array section bound was in a list of bound-types we improperly generated the list of types of the bounds because
getPointeeOrElementType gets the LOWEST level of type (that is, digs
through ALL array types to get to the base-est of types) when what
we really wanted was 1 layer of pointer/array removed.

This patch fixes it and adds a test that showed the problem by re-ordering the existing ones. This wasn't previously obvious by chance, since the 'array-index-only' variants ended up generating the recipe, and not the bounds.

---

Full diff: https://github.com/llvm/llvm-project/pull/162499.diff

2 Files Affected:

• (modified) clang/lib/CIR/CodeGen/CIRGenOpenACC.cpp (+4-1)
• (modified) clang/test/CIR/CodeGenOpenACC/private-clause-array-recipes-int.cpp (+2-2)

diff --git a/clang/lib/CIR/CodeGen/CIRGenOpenACC.cpp b/clang/lib/CIR/CodeGen/CIRGenOpenACC.cpp
index a9af753381db3..4cf2237468afd 100644
--- a/clang/lib/CIR/CodeGen/CIRGenOpenACC.cpp
+++ b/clang/lib/CIR/CodeGen/CIRGenOpenACC.cpp
@@ -87,7 +87,10 @@ CIRGenFunction::getOpenACCDataOperandInfo(const Expr *e) {
     if (const auto *section = dyn_cast<ArraySectionExpr>(curVarExpr)) {
       QualType baseTy = ArraySectionExpr::getBaseOriginalType(
           section->getBase()->IgnoreParenImpCasts());
-      boundTypes.push_back(QualType(baseTy->getPointeeOrArrayElementType(), 0));
+      if (auto *at = getContext().getAsArrayType(baseTy))
+        boundTypes.push_back(at->getElementType());
+      else
+        boundTypes.push_back(baseTy->getPointeeType());
     } else {
       boundTypes.push_back(curVarExpr->getType());
     }
diff --git a/clang/test/CIR/CodeGenOpenACC/private-clause-array-recipes-int.cpp b/clang/test/CIR/CodeGenOpenACC/private-clause-array-recipes-int.cpp
index e83e548cd138b..74cb567076f21 100644
--- a/clang/test/CIR/CodeGenOpenACC/private-clause-array-recipes-int.cpp
+++ b/clang/test/CIR/CodeGenOpenACC/private-clause-array-recipes-int.cpp
@@ -21,7 +21,7 @@ void do_things(unsigned A, unsigned B) {
   ;
 
   T TwoArr[5][5];
-#pragma acc parallel private(TwoArr[B][B])
+#pragma acc parallel private(TwoArr[A:B][A:B])
 // CHECK-NEXT: acc.private.recipe @privatization__Bcnt2__ZTSA5_A5_i : !cir.ptr<!cir.array<!cir.array<!s32i x 5> x 5>> init {
 // CHECK-NEXT: ^bb0(%arg0: !cir.ptr<!cir.array<!cir.array<!s32i x 5> x 5>> {{.*}}, %[[BOUND1:.*]]: !acc.data_bounds_ty {{.*}}, %[[BOUND2:.*]]: !acc.data_bounds_ty {{.*}}):
 // CHECK-NEXT: %[[TL_ALLOCA:.*]] = cir.alloca !cir.array<!cir.array<!s32i x 5> x 5>, !cir.ptr<!cir.array<!cir.array<!s32i x 5> x 5>>, ["openacc.private.init"] {alignment = 4 : i64}
@@ -30,7 +30,7 @@ void do_things(unsigned A, unsigned B) {
   ;
 #pragma acc parallel private(TwoArr[B][A:B])
   ;
-#pragma acc parallel private(TwoArr[A:B][A:B])
+#pragma acc parallel private(TwoArr[B][B])
   ;
 #pragma acc parallel private(TwoArr)
 // CHECK-NEXT: acc.private.recipe @privatization__ZTSA5_A5_i : !cir.ptr<!cir.array<!cir.array<!s32i x 5> x 5>> init {