[OpenMP,C/C++] Handle array reference base-pointers in array sections

  Hi Jakub,
as encountered in cases where a program constructs its own deep-copying
for arrays-of-pointers, e.g:

    #pragma omp target enter data map(to:level->vectors[:N])
    for (i = 0; i < N; i++)
      #pragma omp target enter data map(to:level->vectors[i][:N])

We need to treat the part of the array reference before the array section
as a base-pointer (here 'level->vectors[i]'), providing pointer-attachment behavior.

This patch adds this inside handle_omp_array_sections(), tracing the whole
sequence of array dimensions, creating a whole base-pointer reference
iteratively using build_array_ref(). The conditions are that each of the
"absorbed" dimensions must be length==1, and the final reference must be
of pointer-type (so that pointer attachment makes sense).

There's also a little patch in gimplify_scan_omp_clauses(), to make sure
the array-ref base-pointer goes down the right path.

This case was encountered when working to make 534.hpgmgfv_t from
SPEChpc 2021 properly compile. Tested without regressions on trunk.
Okay to go in once stage1 opens?

Thanks,
Chung-Lin

2022-02-21  Chung-Lin Tang  <cltang@codesourcery.com>

gcc/c/ChangeLog:

	* c-typeck.cc (handle_omp_array_sections): Add handling for
	creating array-reference base-pointer attachment clause.

gcc/cp/ChangeLog:

	* semantics.cc (handle_omp_array_sections): Add handling for
	creating array-reference base-pointer attachment clause.

gcc/ChangeLog:

	* gimplify.cc (gimplify_scan_omp_clauses): Add case for
	attach/detach map kind for ARRAY_REF of POINTER_TYPE.

gcc/testsuite/ChangeLog:

	* c-c++-common/gomp/target-enter-data-1.c: Adjust testcase.

libgomp/testsuite/ChangeLog:

	* libgomp.c-c++-common/ptr-attach-2.c: New test.
  

On Mon, Feb 21, 2022 at 11:18:57PM +0800, Chung-Lin Tang wrote:
> as encountered in cases where a program constructs its own deep-copying
> for arrays-of-pointers, e.g:
> 
>    #pragma omp target enter data map(to:level->vectors[:N])
>    for (i = 0; i < N; i++)
>      #pragma omp target enter data map(to:level->vectors[i][:N])
> 
> We need to treat the part of the array reference before the array section
> as a base-pointer (here 'level->vectors[i]'), providing pointer-attachment behavior.
> 
> This patch adds this inside handle_omp_array_sections(), tracing the whole
> sequence of array dimensions, creating a whole base-pointer reference
> iteratively using build_array_ref(). The conditions are that each of the
> "absorbed" dimensions must be length==1, and the final reference must be
> of pointer-type (so that pointer attachment makes sense).
> 
> There's also a little patch in gimplify_scan_omp_clauses(), to make sure
> the array-ref base-pointer goes down the right path.
> 
> This case was encountered when working to make 534.hpgmgfv_t from
> SPEChpc 2021 properly compile. Tested without regressions on trunk.
> Okay to go in once stage1 opens?

I'm afraid this is going in the wrong direction.  The OpenMP 5.0 change
that:
"A locator list item is any lvalue expression, including variables, or an array
section."
is much more general than just allowing -> in the expressions, there can be
function calls and many other.  So, the more code like this patch we add,
the more we'll need to throw away again.  And as it is in OpenMP 5.0, we
really need to throw it away in the GCC 13 cycle.

So, what we really need is add OMP_ARRAY_SECTION tree code, some parser flag
to know that we are inside of an OpenMP clause that allows array sections,
and just where we normally parse ARRAY_REFs if that flag is on also parse
array sections and parse the map/to/from clauses just as normal expressions.
At least for the C FE maybe we'll need to arrange for less folding to be
done because C still folds too much stuff prematurely.
Then when finishing clauses verify that OMP_ARRAY_SECTION trees appear only
where we allow them and not elsewhere (say
foo (1, 2, 3)[:36]
would be ok if foo returns a pointer, but
foo (ptr[0:13], 2, 3)
would not) and then need to differentiate between the cases listed in the
standard which we handle for each . -> [idx] when starting from a var
(in such a case I vaguely recall there are rules for pointer attachments
etc.) or other arbitrary expressions (in that case we just evaluate those
expressions and e.g. in the foo (1, 2, 3)[:36] case basically do
tmp = foo (1, 2, 3);
and mapping of tmp[:36].

> 2022-02-21  Chung-Lin Tang  <cltang@codesourcery.com>
> 
> gcc/c/ChangeLog:
> 
> 	* c-typeck.cc (handle_omp_array_sections): Add handling for
> 	creating array-reference base-pointer attachment clause.
> 
> gcc/cp/ChangeLog:
> 
> 	* semantics.cc (handle_omp_array_sections): Add handling for
> 	creating array-reference base-pointer attachment clause.
> 
> gcc/ChangeLog:
> 
> 	* gimplify.cc (gimplify_scan_omp_clauses): Add case for
> 	attach/detach map kind for ARRAY_REF of POINTER_TYPE.
> 
> gcc/testsuite/ChangeLog:
> 
> 	* c-c++-common/gomp/target-enter-data-1.c: Adjust testcase.
> 
> libgomp/testsuite/ChangeLog:
> 
> 	* libgomp.c-c++-common/ptr-attach-2.c: New test.

	Jakub
  

On Thu, 5 May 2022 10:52:57 +0200
Jakub Jelinek via Gcc-patches <gcc-patches@gcc.gnu.org> wrote:

> On Mon, Feb 21, 2022 at 11:18:57PM +0800, Chung-Lin Tang wrote:
> > as encountered in cases where a program constructs its own
> > deep-copying for arrays-of-pointers, e.g:
> > 
> >    #pragma omp target enter data map(to:level->vectors[:N])
> >    for (i = 0; i < N; i++)
> >      #pragma omp target enter data map(to:level->vectors[i][:N])
> > 
> > We need to treat the part of the array reference before the array
> > section as a base-pointer (here 'level->vectors[i]'), providing
> > pointer-attachment behavior.
> > 
> > This patch adds this inside handle_omp_array_sections(), tracing
> > the whole sequence of array dimensions, creating a whole
> > base-pointer reference iteratively using build_array_ref(). The
> > conditions are that each of the "absorbed" dimensions must be
> > length==1, and the final reference must be of pointer-type (so that
> > pointer attachment makes sense).
> > 
> > There's also a little patch in gimplify_scan_omp_clauses(), to make
> > sure the array-ref base-pointer goes down the right path.
> > 
> > This case was encountered when working to make 534.hpgmgfv_t from
> > SPEChpc 2021 properly compile. Tested without regressions on trunk.
> > Okay to go in once stage1 opens?  
> 
> I'm afraid this is going in the wrong direction.  The OpenMP 5.0
> change that:
> "A locator list item is any lvalue expression, including variables,
> or an array section."
> is much more general than just allowing -> in the expressions, there
> can be function calls and many other.  So, the more code like this
> patch we add, the more we'll need to throw away again.  And as it is
> in OpenMP 5.0, we really need to throw it away in the GCC 13 cycle.
> 
> So, what we really need is add OMP_ARRAY_SECTION tree code, some
> parser flag to know that we are inside of an OpenMP clause that
> allows array sections, and just where we normally parse ARRAY_REFs if
> that flag is on also parse array sections and parse the map/to/from
> clauses just as normal expressions.

All the above (at least) has been done as part of the patch series
posted here:

https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591973.html

> At least for the C FE maybe we'll
> need to arrange for less folding to be done because C still folds too
> much stuff prematurely. Then when finishing clauses verify that
> OMP_ARRAY_SECTION trees appear only where we allow them and not
> elsewhere (say foo (1, 2, 3)[:36]
> would be ok if foo returns a pointer, but
> foo (ptr[0:13], 2, 3)
> would not) and then need to differentiate between the cases listed in
> the standard which we handle for each . -> [idx] when starting from a
> var (in such a case I vaguely recall there are rules for pointer
> attachments etc.) or other arbitrary expressions (in that case we
> just evaluate those expressions and e.g. in the foo (1, 2, 3)[:36]
> case basically do tmp = foo (1, 2, 3);
> and mapping of tmp[:36].

...which also changes/refactors quite a lot regarding how lowering
clauses into mapping nodes works (the "address inspector" bits).
"Weird" cases like mapping the return value from functions doesn't
necessarily DTRT yet -- it wasn't entirely clear how that should/could
work, I don't think.

HTH,

Julian
  

On Thu, May 05, 2022 at 12:46:29PM +0100, Julian Brown wrote:
> All the above (at least) has been done as part of the patch series
> posted here:
> 
> https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591973.html

Ah, ok, so is this patch superseded by that series, or do you want to apply
it just to be removed again?

> > At least for the C FE maybe we'll
> > need to arrange for less folding to be done because C still folds too
> > much stuff prematurely. Then when finishing clauses verify that
> > OMP_ARRAY_SECTION trees appear only where we allow them and not
> > elsewhere (say foo (1, 2, 3)[:36]
> > would be ok if foo returns a pointer, but
> > foo (ptr[0:13], 2, 3)
> > would not) and then need to differentiate between the cases listed in
> > the standard which we handle for each . -> [idx] when starting from a
> > var (in such a case I vaguely recall there are rules for pointer
> > attachments etc.) or other arbitrary expressions (in that case we
> > just evaluate those expressions and e.g. in the foo (1, 2, 3)[:36]
> > case basically do tmp = foo (1, 2, 3);
> > and mapping of tmp[:36].
> 
> ...which also changes/refactors quite a lot regarding how lowering
> clauses into mapping nodes works (the "address inspector" bits).
> "Weird" cases like mapping the return value from functions doesn't
> necessarily DTRT yet -- it wasn't entirely clear how that should/could
> work, I don't think.

I vaguely remember that the ./-/[] handling applies only when it starts
from a variable and as soon as one triggers something else, perhaps
including just *& or similar stuff then only the final lvalue is mapped and
nothing else, but dunno if it is from just discussions about the topic or
what is actually written in the spec, will need to look it up.

	Jakub
  

On Thu, 5 May 2022 14:40:38 +0200
Jakub Jelinek via Gcc-patches <gcc-patches@gcc.gnu.org> wrote:

> On Thu, May 05, 2022 at 12:46:29PM +0100, Julian Brown wrote:
> > All the above (at least) has been done as part of the patch series
> > posted here:
> > 
> > https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591973.html  
> 
> Ah, ok, so is this patch superseded by that series, or do you want to
> apply it just to be removed again?

Regarding this one (Chung-Lin's, not mine), I'm not sure -- I don't
know if it addresses a problem that is still present with my patch
series applied, nor do I know if my series and this patch have been
tested together. We might need to do some work to integrate the bits,
one way or the other.

Thanks,

Julian