OpenMP: Add ME support for 'omp allocate' stack variables

Call GOMP_alloc/free for 'omp allocate' allocated variables. This is
for C only as C++ and Fortran show a sorry already in the FE. Note that
this only applies to stack variables as the C FE shows a sorry for
static variables.

gcc/ChangeLog:

	* gimplify.cc (gimplify_bind_expr): Call GOMP_alloc/free for
	'omp allocate' variables; move stack cleanup after other
	cleanup.
	(omp_notice_variable): Process original decl when decl
	of the value-expression for a 'omp allocate' variable is passed.
	* omp-low.cc (scan_omp_1_op): Handle 'omp allocate' variables

libgomp/ChangeLog:

	* libgomp.texi (OpenMP 5.1 Impl.): Mark 'omp allocate' as
	implemented for C only.
	* testsuite/libgomp.c/allocate-4.c: New test.
	* testsuite/libgomp.c/allocate-5.c: New test.
	* testsuite/libgomp.c/allocate-6.c: New test.

gcc/testsuite/ChangeLog:

	* c-c++-common/gomp/allocate-11.c: Remove C-only dg-message
	for 'sorry, unimplemented'.
	* c-c++-common/gomp/allocate-12.c: Likewise.
	* c-c++-common/gomp/allocate-15.c: Likewise.
	* c-c++-common/gomp/allocate-9.c: Likewise.
	* c-c++-common/gomp/allocate-10.c: New test.
	* c-c++-common/gomp/allocate-17.c: New test.

8 files changed, 201 insertions, 33 deletions

diff --git a/gcc/gimplify.cc b/gcc/gimplify.cc
index a0e8cc2..9f4722f 100644
--- a/gcc/gimplify.cc
+++ b/gcc/gimplify.cc
@@ -36,6 +36,7 @@ along with GCC; see the file COPYING3.  If not see
 #include "cgraph.h"
 #include "tree-pretty-print.h"
 #include "diagnostic-core.h"
+#include "diagnostic.h"		/* For errorcount.  */
 #include "alias.h"
 #include "fold-const.h"
 #include "calls.h"
@@ -1372,6 +1373,7 @@ gimplify_bind_expr (tree *expr_p, gimple_seq *pre_p)
 	      && (attr = lookup_attribute ("omp allocate", DECL_ATTRIBUTES (t)))
 		 != NULL_TREE)
 	    {
+	      gcc_assert (!DECL_HAS_VALUE_EXPR_P (t));
 	      tree alloc = TREE_PURPOSE (TREE_VALUE (attr));
 	      tree align = TREE_VALUE (TREE_VALUE (attr));
 	      /* Allocate directives that appear in a target region must specify
@@ -1396,12 +1398,56 @@ gimplify_bind_expr (tree *expr_p, gimple_seq *pre_p)
 		error_at (DECL_SOURCE_LOCATION (t),
 			  "%<allocate%> directive for %qD inside a target "
 			  "region must specify an %<allocator%> clause", t);
-	      else if (align != NULL_TREE
-		       || alloc == NULL_TREE
-		       || !integer_onep (alloc))
-	        sorry_at (DECL_SOURCE_LOCATION (t),
-			  "OpenMP %<allocate%> directive, used for %qD, not "
-			  "yet supported", t);
+	      /* Skip for omp_default_mem_alloc (= 1),
+		 unless align is present. */
+	      else if (!errorcount
+		       && (align != NULL_TREE
+			   || alloc == NULL_TREE
+			   || !integer_onep (alloc)))
+		{
+		  tree tmp = build_pointer_type (TREE_TYPE (t));
+		  tree v = create_tmp_var (tmp, get_name (t));
+		  DECL_IGNORED_P (v) = 0;
+		  tmp = remove_attribute ("omp allocate", DECL_ATTRIBUTES (t));
+		  DECL_ATTRIBUTES (v)
+		    = tree_cons (get_identifier ("omp allocate var"),
+				 build_tree_list (NULL_TREE, t), tmp);
+		  tmp = build_fold_indirect_ref (v);
+		  TREE_THIS_NOTRAP (tmp) = 1;
+		  SET_DECL_VALUE_EXPR (t, tmp);
+		  DECL_HAS_VALUE_EXPR_P (t) = 1;
+		  tree sz = TYPE_SIZE_UNIT (TREE_TYPE (t));
+		  if (alloc == NULL_TREE)
+		    alloc = build_zero_cst (ptr_type_node);
+		  if (align == NULL_TREE)
+		    align = build_int_cst (size_type_node, DECL_ALIGN_UNIT (t));
+		  else
+		    align = build_int_cst (size_type_node,
+					   MAX (tree_to_uhwi (align),
+						DECL_ALIGN_UNIT (t)));
+		  tmp = builtin_decl_explicit (BUILT_IN_GOMP_ALLOC);
+		  tmp = build_call_expr_loc (DECL_SOURCE_LOCATION (t), tmp,
+					     3, align, sz, alloc);
+		  tmp = fold_build2_loc (DECL_SOURCE_LOCATION (t), MODIFY_EXPR,
+					 TREE_TYPE (v), v,
+					 fold_convert (TREE_TYPE (v), tmp));
+		  gcc_assert (BIND_EXPR_BODY (bind_expr) != NULL_TREE
+			      && (TREE_CODE (BIND_EXPR_BODY (bind_expr))
+				  == STATEMENT_LIST));
+		  tree_stmt_iterator e = tsi_start (BIND_EXPR_BODY (bind_expr));
+		  while (!tsi_end_p (e))
+		    {
+		      if ((TREE_CODE (*e) == DECL_EXPR
+			   && TREE_OPERAND (*e, 0) == t)
+			  || (TREE_CODE (*e) == CLEANUP_POINT_EXPR
+			      && TREE_CODE (TREE_OPERAND (*e, 0)) == DECL_EXPR
+			      && TREE_OPERAND (TREE_OPERAND (*e, 0), 0) == t))
+		      break;
+		      ++e;
+		    }
+		  gcc_assert (!tsi_end_p (e));
+		  tsi_link_before (&e, tmp, TSI_SAME_STMT);
+		}
 	    }
 
 	  /* Mark variable as local.  */
@@ -1486,22 +1532,6 @@ gimplify_bind_expr (tree *expr_p, gimple_seq *pre_p)
   cleanup = NULL;
   stack_save = NULL;
 
-  /* If the code both contains VLAs and calls alloca, then we cannot reclaim
-     the stack space allocated to the VLAs.  */
-  if (gimplify_ctxp->save_stack && !gimplify_ctxp->keep_stack)
-    {
-      gcall *stack_restore;
-
-      /* Save stack on entry and restore it on exit.  Add a try_finally
-	 block to achieve this.  */
-      build_stack_save_restore (&stack_save, &stack_restore);
-
-      gimple_set_location (stack_save, start_locus);
-      gimple_set_location (stack_restore, end_locus);
-
-      gimplify_seq_add_stmt (&cleanup, stack_restore);
-    }
-
   /* Add clobbers for all variables that go out of scope.  */
   for (t = BIND_EXPR_VARS (bind_expr); t ; t = DECL_CHAIN (t))
     {
@@ -1509,6 +1539,17 @@ gimplify_bind_expr (tree *expr_p, gimple_seq *pre_p)
 	  && !is_global_var (t)
 	  && DECL_CONTEXT (t) == current_function_decl)
 	{
+	  if (flag_openmp
+	      && DECL_HAS_VALUE_EXPR_P (t)
+	      && TREE_USED (t)
+	      && lookup_attribute ("omp allocate", DECL_ATTRIBUTES (t)))
+	    {
+	      tree tmp = builtin_decl_explicit (BUILT_IN_GOMP_FREE);
+	      tmp = build_call_expr_loc (end_locus, tmp, 2,
+					 TREE_OPERAND (DECL_VALUE_EXPR (t), 0),
+					 build_zero_cst (ptr_type_node));
+	      gimplify_and_add (tmp, &cleanup);
+	    }
 	  if (!DECL_HARD_REGISTER (t)
 	      && !TREE_THIS_VOLATILE (t)
 	      && !DECL_HAS_VALUE_EXPR_P (t)
@@ -1565,6 +1606,22 @@ gimplify_bind_expr (tree *expr_p, gimple_seq *pre_p)
 	gimplify_ctxp->live_switch_vars->remove (t);
     }
 
+  /* If the code both contains VLAs and calls alloca, then we cannot reclaim
+     the stack space allocated to the VLAs.  */
+  if (gimplify_ctxp->save_stack && !gimplify_ctxp->keep_stack)
+    {
+      gcall *stack_restore;
+
+      /* Save stack on entry and restore it on exit.  Add a try_finally
+	 block to achieve this.  */
+      build_stack_save_restore (&stack_save, &stack_restore);
+
+      gimple_set_location (stack_save, start_locus);
+      gimple_set_location (stack_restore, end_locus);
+
+      gimplify_seq_add_stmt (&cleanup, stack_restore);
+    }
+
   if (ret_clauses)
     {
       gomp_target *stmt;
@@ -7894,6 +7951,13 @@ omp_notice_variable (struct gimplify_omp_ctx *ctx, tree decl, bool in_code)
   if (error_operand_p (decl))
     return false;
 
+  if (DECL_ARTIFICIAL (decl))
+    {
+      tree attr = lookup_attribute ("omp allocate var", DECL_ATTRIBUTES (decl));
+      if (attr)
+	decl = TREE_VALUE (TREE_VALUE (attr));
+    }
+
   if (ctx->region_type == ORT_NONE)
     return lang_hooks.decls.omp_disregard_value_expr (decl, false);
 
diff --git a/gcc/omp-low.cc b/gcc/omp-low.cc
index 5d7c32d..b0c3ef7 100644
--- a/gcc/omp-low.cc
+++ b/gcc/omp-low.cc
@@ -3951,6 +3951,7 @@ scan_omp_1_op (tree *tp, int *walk_subtrees, void *data)
   struct walk_stmt_info *wi = (struct walk_stmt_info *) data;
   omp_context *ctx = (omp_context *) wi->info;
   tree t = *tp;
+  tree tmp;
 
   switch (TREE_CODE (t))
     {
@@ -3960,12 +3961,37 @@ scan_omp_1_op (tree *tp, int *walk_subtrees, void *data)
     case RESULT_DECL:
       if (ctx)
 	{
+	  tmp = NULL_TREE;
+	  if (TREE_CODE (t) == VAR_DECL
+	      && (tmp = lookup_attribute ("omp allocate var",
+					  DECL_ATTRIBUTES (t))) != NULL_TREE)
+	    t = TREE_VALUE (TREE_VALUE (tmp));
 	  tree repl = remap_decl (t, &ctx->cb);
 	  gcc_checking_assert (TREE_CODE (repl) != ERROR_MARK);
-	  *tp = repl;
+	  if (tmp != NULL_TREE  && t != repl)
+	    *tp = build_fold_addr_expr (repl);
+	  else if (tmp == NULL_TREE)
+	    *tp = repl;
 	}
       break;
 
+    case INDIRECT_REF:
+    case MEM_REF:
+      if (ctx
+	  && TREE_CODE (TREE_OPERAND (t, 0)) == VAR_DECL
+	  && ((tmp = lookup_attribute ("omp allocate var",
+				       DECL_ATTRIBUTES (TREE_OPERAND (t, 0))))
+	       != NULL_TREE))
+	{
+	  tmp = TREE_VALUE (TREE_VALUE (tmp));
+	  tree repl = remap_decl (tmp, &ctx->cb);
+	  gcc_checking_assert (TREE_CODE (repl) != ERROR_MARK);
+	  if (tmp != repl)
+	    *tp = repl;
+	  break;
+	}
+      gcc_fallthrough ();
+
     default:
       if (ctx && TYPE_P (t))
 	*tp = remap_type (t, &ctx->cb);
diff --git a/gcc/testsuite/c-c++-common/gomp/allocate-10.c b/gcc/testsuite/c-c++-common/gomp/allocate-10.c
new file mode 100644
index 0000000..7e8f579
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/allocate-10.c
@@ -0,0 +1,49 @@
+/* TODO: enable for C++ once implemented. */
+/* { dg-do compile { target c } } */
+/* { dg-additional-options "-Wall -fdump-tree-gimple" } */
+
+typedef enum omp_allocator_handle_t
+#if __cplusplus >= 201103L
+: __UINTPTR_TYPE__
+#endif
+{
+  omp_default_mem_alloc = 1,
+  __omp_allocator_handle_t_max__ = __UINTPTR_MAX__
+} omp_allocator_handle_t;
+
+void
+f()
+{
+  int n;
+  int A[n]; /* { dg-warning "'n' is used uninitialized" } */
+  /* { dg-warning "unused variable 'A'" "" { target *-*-* } .-1 } */
+}
+
+void
+h1()
+{
+  omp_allocator_handle_t my_handle;
+  int B1[3]; /* { dg-warning "'my_handle' is used uninitialized" } */
+  /* { dg-warning "variable 'B1' set but not used" "" { target *-*-* } .-1 } */
+  #pragma omp allocate(B1) allocator(my_handle)
+  B1[0] = 5;
+  /* { dg-final { scan-tree-dump-times "__builtin_GOMP_alloc" 1 "gimple" } } */
+  /* { dg-final { scan-tree-dump-times "B1.\[0-9\]+ = __builtin_GOMP_alloc \\(4, 12, my_handle\\);" 1 "gimple" } } */
+  /* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(B1.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+}
+
+void
+h2()
+{
+  omp_allocator_handle_t my_handle;
+  int B2[3];  /* { dg-warning "unused variable 'B2'" } */
+  #pragma omp allocate(B2) allocator(my_handle) /* No warning as 'B2' is unused */
+}
+
+void
+h3()
+{
+  omp_allocator_handle_t my_handle;
+  int B3[3] = {1,2,3};  /* { dg-warning "unused variable 'B3'" } */
+  #pragma omp allocate(B3) allocator(my_handle) /* No warning as 'B3' is unused */
+}
diff --git a/gcc/testsuite/c-c++-common/gomp/allocate-11.c b/gcc/testsuite/c-c++-common/gomp/allocate-11.c
index f9ad50a..dceb97f 100644
--- a/gcc/testsuite/c-c++-common/gomp/allocate-11.c
+++ b/gcc/testsuite/c-c++-common/gomp/allocate-11.c
@@ -10,7 +10,6 @@ f (int i)
   switch (i)  /* { dg-note "switch starts here" } */
     {
       int j;  /* { dg-note "'j' declared here" } */
-      /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target *-*-* } .-1 } */
       #pragma omp allocate(j)
     case 42:  /* { dg-error "switch jumps over OpenMP 'allocate' allocation" } */
       bar ();
@@ -30,9 +29,7 @@ h (int i2)
   return 5;
 
   int k2;  /* { dg-note "'k2' declared here" } */
-  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target *-*-* } .-1 } */
   int j2 = 4;  /* { dg-note "'j2' declared here" } */
-  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target *-*-* } .-1 } */
   #pragma omp allocate(k2, j2)
 label:  /* { dg-note "label 'label' defined here" } */
   k2 = 4;
diff --git a/gcc/testsuite/c-c++-common/gomp/allocate-12.c b/gcc/testsuite/c-c++-common/gomp/allocate-12.c
index 3c7c3bb..1b77db9 100644
--- a/gcc/testsuite/c-c++-common/gomp/allocate-12.c
+++ b/gcc/testsuite/c-c++-common/gomp/allocate-12.c
@@ -17,7 +17,6 @@ f ()
   omp_allocator_handle_t my_allocator;
   int n = 5;  /* { dg-note "to be allocated variable declared here" } */
   my_allocator = omp_default_mem_alloc; /* { dg-note "modified here" } */
-  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target *-*-* } .-2 } */
   #pragma omp allocate(n) allocator(my_allocator)  /* { dg-error "variable 'my_allocator' used in the 'allocator' clause must not be modified between declaration of 'n' and its 'allocate' directive" } */
   n = 7;
   return n;
@@ -28,7 +27,6 @@ int
 g ()
 {
   int n = 5;  /* { dg-note "to be allocated variable declared here" } */
-  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target *-*-* } .-1 } */
   omp_allocator_handle_t my_allocator = omp_low_lat_mem_alloc;  /* { dg-note "declared here" } */
   #pragma omp allocate(n) allocator(my_allocator)  /* { dg-error "variable 'my_allocator' used in the 'allocator' clause must be declared before 'n'" } */
   n = 7;
@@ -42,7 +40,6 @@ h ()
      see gomp/allocate-10.c.  */
   omp_allocator_handle_t my_allocator;
   int n = 5;
-  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target *-*-* } .-1 } */
   #pragma omp allocate(n) allocator(my_allocator)
   n = 7;
   return n;
diff --git a/gcc/testsuite/c-c++-common/gomp/allocate-15.c b/gcc/testsuite/c-c++-common/gomp/allocate-15.c
index d9600f9..15105b91 100644
--- a/gcc/testsuite/c-c++-common/gomp/allocate-15.c
+++ b/gcc/testsuite/c-c++-common/gomp/allocate-15.c
@@ -8,7 +8,7 @@ void
 f ()
 {
 
-  int var;  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive, used for 'var', not yet supported" } */
+  int var;
   #pragma omp allocate(var)
   var = 5;
 }
@@ -21,7 +21,7 @@ h ()
    #pragma omp parallel
     #pragma omp serial
      {
-       int var2[5];  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive, used for 'var2', not yet supported" } */
+       int var2[5];
        #pragma omp allocate(var2)
        var2[0] = 7;
      }
diff --git a/gcc/testsuite/c-c++-common/gomp/allocate-17.c b/gcc/testsuite/c-c++-common/gomp/allocate-17.c
new file mode 100644
index 0000000..f75af0c
--- /dev/null
+++ b/gcc/testsuite/c-c++-common/gomp/allocate-17.c
@@ -0,0 +1,37 @@
+/* This file has a syntax error but should not ICE.
+   Namely, a '}' is missing in one(). */
+
+typedef enum omp_allocator_handle_t
+#if __cplusplus >= 201103L
+: __UINTPTR_TYPE__
+#endif
+{
+  omp_default_mem_alloc = 1,
+  omp_low_lat_mem_alloc = 5,
+  __omp_allocator_handle_t_max__ = __UINTPTR_MAX__
+} omp_allocator_handle_t;
+
+#include <stdint.h>
+
+void
+one ()
+{  /* { dg-note "to match this '\{'" "" { target c++ } } */
+  int result = 0, n = 3;
+  #pragma omp target map(tofrom: result) firstprivate(n)
+    {
+      int var = 5; //, var2[n];
+      #pragma omp allocate(var) align(128) allocator(omp_low_lat_mem_alloc) /* { dg-message "sorry, unimplemented: '#pragma omp allocate' not yet supported" "" { target c++ } } */
+       var = 7;
+}
+
+void
+two ()
+{ /* { dg-error "a function-definition is not allowed here before '\{' token" "" { target c++ } } */
+  int scalar = 44;
+  #pragma omp allocate(scalar)
+
+  #pragma omp parallel firstprivate(scalar)
+    scalar = 33;
+}
+/* { dg-error "expected declaration or statement at end of input" "" { target c } .-1 } */
+/* { dg-error "expected '\}' at end of input" "" { target c++ } .-2 } */
diff --git a/gcc/testsuite/c-c++-common/gomp/allocate-9.c b/gcc/testsuite/c-c++-common/gomp/allocate-9.c
index 8e01041..3c11080 100644
--- a/gcc/testsuite/c-c++-common/gomp/allocate-9.c
+++ b/gcc/testsuite/c-c++-common/gomp/allocate-9.c
@@ -86,8 +86,6 @@ int g()
 /* { dg-note "declared here" "" { target c } .-8 } */
 /* { dg-message "sorry, unimplemented: '#pragma omp allocate' not yet supported" "" { target c++ } .-2 } */
     return c2+a2+b2;
-  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target c } .-5 } */
-  /* { dg-message "sorry, unimplemented: OpenMP 'allocate' directive" "" { target c } .-12 } */
   }
 }