[PATCH, OpenMP] PR103642 - Fix omp-low ICE for indirect references based off component access

[PATCH, OpenMP] PR103642 - Fix omp-low ICE for indirect references based off component access

[Chung-Lin Tang]

[-- 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;
+}

Re: [PATCH, OpenMP] PR103642 - Fix omp-low ICE for indirect references based off component access

[Chung-Lin Tang]

Ping.

On 2022/1/3 10:15 PM, Chung-Lin Tang wrote:
> 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.
> 
> 
> 
> 
> 

Re: [PATCH, OpenMP] PR103642 - Fix omp-low ICE for indirect references based off component access

[Jakub Jelinek]

On Mon, Jan 03, 2022 at 10:15:26PM +0800, Chung-Lin Tang wrote:
> 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.

Ok, with a small nit:

> +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;
> +  }

Please free the pointers afterwards.

	Jakub