[committed,openacc] Fix SIGSEGV in oacc_entry_exit_ok_1

Message ID 20190616074939.GA32295@delia
State New
Headers show
Series
  • [committed,openacc] Fix SIGSEGV in oacc_entry_exit_ok_1
Related show

Commit Message

Tom de Vries June 16, 2019, 7:49 a.m.
Hi,

When compiling the test-case with r268755, we run into a SIGSEGV in
oacc_entry_exit_ok_1 when trying to dereference a NULL red:
...
                      struct reduction_info *red;
                      red = reduction_phi (reduction_list, use_stmt);
                      tree val = PHI_RESULT (red->keep_res);
...

Fix this by handling ref == NULL.

Bootstrapped and reg-tested on x86_64.
Build and reg-tested on x86_64 with nvptx accelerator.

Committed to trunk.

Thanks,
- Tom

[openacc] Fix SIGSEGV in oacc_entry_exit_ok_1

2019-06-16  Tom de Vries  <tdevries@suse.de>

	PR tree-optimization/89376
	* tree-parloops.c (oacc_entry_exit_ok_1): Handle red == NULL.

	* testsuite/libgomp.oacc-c-c++-common/pr89376.c: New test.

---
 gcc/tree-parloops.c                                   |  6 +++---
 libgomp/testsuite/libgomp.oacc-c-c++-common/pr89376.c | 15 +++++++++++++++
 2 files changed, 18 insertions(+), 3 deletions(-)

Patch

diff --git a/gcc/tree-parloops.c b/gcc/tree-parloops.c
index 9de154e722f..6b8c8cd5b75 100644
--- a/gcc/tree-parloops.c
+++ b/gcc/tree-parloops.c
@@ -3052,11 +3052,11 @@  oacc_entry_exit_ok_1 (bitmap in_loop_bbs, vec<basic_block> region_bbs,
 		{
 		  use_operand_p use_p;
 		  gimple *use_stmt;
+		  struct reduction_info *red;
 		  single_imm_use (lhs, &use_p, &use_stmt);
-		  if (gimple_code (use_stmt) == GIMPLE_PHI)
+		  if (gimple_code (use_stmt) == GIMPLE_PHI
+		      && (red = reduction_phi (reduction_list, use_stmt)))
 		    {
-		      struct reduction_info *red;
-		      red = reduction_phi (reduction_list, use_stmt);
 		      tree val = PHI_RESULT (red->keep_res);
 		      if (has_single_use (val))
 			{
diff --git a/libgomp/testsuite/libgomp.oacc-c-c++-common/pr89376.c b/libgomp/testsuite/libgomp.oacc-c-c++-common/pr89376.c
new file mode 100644
index 00000000000..7f6c83254d5
--- /dev/null
+++ b/libgomp/testsuite/libgomp.oacc-c-c++-common/pr89376.c
@@ -0,0 +1,15 @@ 
+/* { dg-do run } */
+/* { dg-additional-options "-fno-tree-ch -fno-tree-dce -fno-tree-vrp" } */
+
+int
+main (void)
+{
+  int fa;
+
+  #pragma acc kernels
+  for (int rw = 0; rw < 1; ++rw)
+    fa = 0;
+
+  return 0;
+}
+