From: Chung-Lin Tang <cltang@codesourcery.com>
To: gcc-patches <gcc-patches@gcc.gnu.org>,
Jakub Jelinek <jakub@redhat.com>,
Tobias Burnus <tobias@codesourcery.com>
Subject: [PATCH, OpenMP] PR103642 - Fix omp-low ICE for indirect references based off component access
Date: Mon, 3 Jan 2022 22:15:26 +0800 [thread overview]
Message-ID: <7233befe-4c44-b275-c4ac-82af937f32f0@codesourcery.com> (raw)
[-- Attachment #1: Type: text/plain, Size: 1100 bytes --]
This issue was triggered after the patch extending syntax for component access
in map clauses
(https://gcc.gnu.org/git/gitweb.cgi?p=gcc.git;h=0ab29cf0bb68960c)
In gimplify_scan_omp_clauses, the case for handling indirect accesses (which creates
firstprivate ptr and zero-length array section map for such decls) was erroneously
went into for non-pointer cases (here being the base struct decl), so added the
appropriate checks there.
Added new testcase is a compile only test for the ICE. The original omptests t-partial-struct
test actually should not execute correctly, because for map(t.s->a[:N]), map(t.s[:1])
is not implicitly mapped, thus the entire offloaded access does not work as is.
(fixing that omptests test is out of scope here)
Tested without regressions, okay for trunk?
Thanks,
Chung-Lin
2022-01-03 Chung-Lin Tang <cltang@codesourcery.com>
gcc/ChangeLog:
PR middle-end/103642
* gimplify.c (gimplify_scan_omp_clauses): Do not do indir_p handling
for non-pointer or non-reference-to-pointer cases.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/pr103642.c: New test.
[-- Attachment #2: pr103642.patch --]
[-- Type: text/plain, Size: 1242 bytes --]
diff --git a/gcc/gimplify.c b/gcc/gimplify.c
index b118c72f62c..bdc8189c2a7 100644
--- a/gcc/gimplify.c
+++ b/gcc/gimplify.c
@@ -9543,7 +9543,10 @@ gimplify_scan_omp_clauses (tree *list_p, gimple_seq *pre_p,
== REFERENCE_TYPE))
decl = TREE_OPERAND (decl, 0);
}
- if (decl != orig_decl && DECL_P (decl) && indir_p)
+ if (decl != orig_decl && DECL_P (decl) && indir_p
+ && (TREE_CODE (TREE_TYPE (decl)) == POINTER_TYPE
+ || (decl_ref
+ && TREE_CODE (TREE_TYPE (decl_ref)) == POINTER_TYPE)))
{
gomp_map_kind k
= ((code == OACC_EXIT_DATA || code == OMP_TARGET_EXIT_DATA)
diff --git a/gcc/testsuite/c-c++-common/gomp/pr103642.c b/gcc/testsuite/c-c++-common/gomp/pr103642.c
new file mode 100644
index 00000000000..c5451596b69
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/pr103642.c
@@ -0,0 +1,31 @@
+/* PR middle-end/103642 */
+/* { dg-do compile } */
+
+#include <stdlib.h>
+
+typedef struct
+{
+ int *a;
+} S;
+
+typedef struct
+{
+ S *s;
+ int *ptr;
+} T;
+
+#define N 10
+
+int main (void)
+{
+ T t;
+ t.s = (S *) malloc (sizeof (S));
+ t.s->a = (int *) malloc (sizeof(int) * N);
+
+ #pragma omp target map(from: t.s->a[:N])
+ {
+ t.s->a[0] = 1;
+ }
+
+ return 0;
+}
next reply other threads:[~2022-01-03 14:15 UTC|newest]
Thread overview: 3+ messages / expand[flat|nested] mbox.gz Atom feed top
2022-01-03 14:15 Chung-Lin Tang [this message]
2022-01-17 14:55 ` Chung-Lin Tang
2022-01-18 14:09 ` Jakub Jelinek
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=7233befe-4c44-b275-c4ac-82af937f32f0@codesourcery.com \
--to=cltang@codesourcery.com \
--cc=gcc-patches@gcc.gnu.org \
--cc=jakub@redhat.com \
--cc=tobias@codesourcery.com \
/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).