public inbox for gcc-patches@gcc.gnu.org
 help / color / mirror / Atom feed
From: Jakub Jelinek <jakub@redhat.com>
To: Chung-Lin Tang <cltang@codesourcery.com>
Cc: gcc-patches <gcc-patches@gcc.gnu.org>
Subject: Re: [PATCH, OpenMP, C/C++] Handle array reference base-pointers in array sections
Date: Thu, 5 May 2022 10:52:57 +0200	[thread overview]
Message-ID: <YnOQadPHUl+/pWUK@tucnak> (raw)
In-Reply-To: <65e1462e-d17f-2975-1401-358fe9c69e28@codesourcery.com>

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


  reply	other threads:[~2022-05-05  8:53 UTC|newest]

Thread overview: 6+ messages / expand[flat|nested]  mbox.gz  Atom feed  top
2022-02-21 15:18 Chung-Lin Tang
2022-05-05  8:52 ` Jakub Jelinek [this message]
2022-05-05 11:46   ` Julian Brown
2022-05-05 11:46     ` Julian Brown
2022-05-05 12:40     ` Jakub Jelinek
2022-05-05 15:59       ` Julian Brown

Reply instructions:

You may reply publicly to this message via plain-text email
using any one of the following methods:

* Save the following mbox file, import it into your mail client,
  and reply-to-all from there: mbox

  Avoid top-posting and favor interleaved quoting:
  https://en.wikipedia.org/wiki/Posting_style#Interleaved_style

* Reply using the --to, --cc, and --in-reply-to
  switches of git-send-email(1):

  git send-email \
    --in-reply-to=YnOQadPHUl+/pWUK@tucnak \
    --to=jakub@redhat.com \
    --cc=cltang@codesourcery.com \
    --cc=gcc-patches@gcc.gnu.org \
    /path/to/YOUR_REPLY

  https://kernel.org/pub/software/scm/git/docs/git-send-email.html

* If your mail client supports setting the In-Reply-To header
  via mailto: links, try the mailto: link
Be sure your reply has a Subject: header at the top and a blank line before the message body.
This is a public inbox, see mirroring instructions
for how to clone and mirror all data and code used for this inbox;
as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).