public inbox for gcc-cvs@sourceware.org
help / color / mirror / Atom feed
* [gcc r14-4176] OpenMP: Add ME support for 'omp allocate' stack variables
@ 2023-09-20 14:04 Tobias Burnus
0 siblings, 0 replies; only message in thread
From: Tobias Burnus @ 2023-09-20 14:04 UTC (permalink / raw)
To: gcc-cvs
https://gcc.gnu.org/g:1a554a2c9f33fdb3c170f1c37274037ece050114
commit r14-4176-g1a554a2c9f33fdb3c170f1c37274037ece050114
Author: Tobias Burnus <tobias@codesourcery.com>
Date: Wed Sep 20 16:03:19 2023 +0200
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.
Diff:
---
gcc/gimplify.cc | 108 +++++++--
gcc/omp-low.cc | 28 ++-
gcc/testsuite/c-c++-common/gomp/allocate-10.c | 49 ++++
gcc/testsuite/c-c++-common/gomp/allocate-11.c | 3 -
gcc/testsuite/c-c++-common/gomp/allocate-12.c | 3 -
gcc/testsuite/c-c++-common/gomp/allocate-15.c | 4 +-
gcc/testsuite/c-c++-common/gomp/allocate-17.c | 37 +++
gcc/testsuite/c-c++-common/gomp/allocate-9.c | 2 -
libgomp/libgomp.texi | 5 +-
libgomp/testsuite/libgomp.c/allocate-4.c | 84 +++++++
libgomp/testsuite/libgomp.c/allocate-5.c | 126 ++++++++++
libgomp/testsuite/libgomp.c/allocate-6.c | 319 ++++++++++++++++++++++++++
12 files changed, 733 insertions(+), 35 deletions(-)
diff --git a/gcc/gimplify.cc b/gcc/gimplify.cc
index a0e8cc2199d..9f4722f7458 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 5d7c32dac39..b0c3ef7a9cc 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 00000000000..7e8f579871c
--- /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 f9ad50abb7f..dceb97f8c5f 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 3c7c3bb3a2b..1b77db9bd6f 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 d9600f96c46..15105b9102e 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 00000000000..f75af0c2d93
--- /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 8e010419a5f..3c11080dd16 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 } */
}
}
diff --git a/libgomp/libgomp.texi b/libgomp/libgomp.texi
index f5cb5b643a2..f21557c3c52 100644
--- a/libgomp/libgomp.texi
+++ b/libgomp/libgomp.texi
@@ -225,7 +225,7 @@ The OpenMP 4.5 specification is fully supported.
@item Predefined memory spaces, memory allocators, allocator traits
@tab Y @tab See also @ref{Memory allocation}
@item Memory management routines @tab Y @tab
-@item @code{allocate} directive @tab N @tab
+@item @code{allocate} directive @tab P @tab Only C, only stack variables
@item @code{allocate} clause @tab P @tab Initial support
@item @code{use_device_addr} clause on @code{target data} @tab Y @tab
@item @code{ancestor} modifier on @code{device} clause @tab Y @tab
@@ -296,7 +296,8 @@ The OpenMP 4.5 specification is fully supported.
@item Loop transformation constructs @tab N @tab
@item @code{strict} modifier in the @code{grainsize} and @code{num_tasks}
clauses of the @code{taskloop} construct @tab Y @tab
-@item @code{align} clause in @code{allocate} directive @tab N @tab
+@item @code{align} clause in @code{allocate} directive @tab P
+ @tab Only C (and only stack variables)
@item @code{align} modifier in @code{allocate} clause @tab Y @tab
@item @code{thread_limit} clause to @code{target} construct @tab Y @tab
@item @code{has_device_addr} clause to @code{target} construct @tab Y @tab
diff --git a/libgomp/testsuite/libgomp.c/allocate-4.c b/libgomp/testsuite/libgomp.c/allocate-4.c
new file mode 100644
index 00000000000..e81cc4093aa
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c/allocate-4.c
@@ -0,0 +1,84 @@
+/* TODO: move to ../libgomp.c-c++-common once C++ is implemented. */
+/* NOTE: { target c } is unsupported with with the C compiler. */
+
+/* { dg-do run } */
+/* { dg-additional-options "-fdump-tree-gimple" } */
+
+#include <omp.h>
+#include <stdint.h>
+
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_alloc \\(" 5 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(" 5 "gimple" } } */
+
+
+int one ()
+{
+ int sum = 0;
+ #pragma omp allocate(sum)
+ /* { dg-final { scan-tree-dump-times "sum\\.\[0-9\]+ = __builtin_GOMP_alloc \\(4, 4, 0B\\);" 1 "gimple" } } */
+ /* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(sum\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+
+ /* NOTE: Initializer cannot be omp_init_allocator - as 'A' is
+ in the same scope and the auto-omp_free comes later than
+ any omp_destroy_allocator. */
+ omp_allocator_handle_t my_allocator = omp_low_lat_mem_alloc;
+ int n = 25;
+ int A[n];
+ #pragma omp allocate(A) align(128) allocator(my_allocator)
+ /* { dg-final { scan-tree-dump-times "A\\.\[0-9\]+ = __builtin_GOMP_alloc \\(128, _\[0-9\]+, my_allocator\\);" 1 "gimple" } } */
+ /* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(A\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+
+ if (((intptr_t)A) % 128 != 0)
+ __builtin_abort ();
+ for (int i = 0; i < n; ++i)
+ A[i] = i;
+
+ omp_alloctrait_t traits[1] = { { omp_atk_alignment, 64 } };
+ my_allocator = omp_init_allocator(omp_low_lat_mem_space,1,traits);
+ {
+ int B[n] = { };
+ int C[5] = {1,2,3,4,5};
+ #pragma omp allocate(B,C) allocator(my_allocator)
+ /* { dg-final { scan-tree-dump-times "B\\.\[0-9\]+ = __builtin_GOMP_alloc \\(\[0-9\]+, _\[0-9\]+, my_allocator\\);" 1 "gimple" } } */
+ /* { dg-final { scan-tree-dump-times "C\\.\[0-9\]+ = __builtin_GOMP_alloc \\(\[0-9\]+, 20, my_allocator\\);" 1 "gimple" } } */
+ /* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(B\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+ /* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(C\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+
+ int D[5] = {11,22,33,44,55};
+ #pragma omp allocate(D) align(256)
+ /* { dg-final { scan-tree-dump-times "D\\.\[0-9\]+ = __builtin_GOMP_alloc \\(256, 20, 0B\\);" 1 "gimple" } } */
+ /* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(D\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+
+ if (((intptr_t) B) % 64 != 0)
+ __builtin_abort ();
+ if (((intptr_t) C) % 64 != 0)
+ __builtin_abort ();
+ if (((intptr_t) D) % 64 != 0)
+ __builtin_abort ();
+
+ for (int i = 0; i < 5; ++i)
+ {
+ if (C[i] != i+1)
+ __builtin_abort ();
+ if (D[i] != i+1 + 10*(i+1))
+ __builtin_abort ();
+ }
+
+ for (int i = 0; i < n; ++i)
+ {
+ if (B[i] != 0)
+ __builtin_abort ();
+ sum += A[i]+B[i]+C[i%5]+D[i%5];
+ }
+ }
+ omp_destroy_allocator (my_allocator);
+ return sum;
+}
+
+int
+main ()
+{
+ if (one () != 1200)
+ __builtin_abort ();
+ return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c/allocate-5.c b/libgomp/testsuite/libgomp.c/allocate-5.c
new file mode 100644
index 00000000000..beaf16440e1
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c/allocate-5.c
@@ -0,0 +1,126 @@
+/* TODO: move to ../libgomp.c-c++-common once C++ is implemented. */
+/* NOTE: { target c } is unsupported with with the C compiler. */
+
+/* { dg-do run } */
+/* { dg-additional-options "-fdump-tree-gimple" } */
+
+#include <omp.h>
+#include <stdint.h>
+
+/* { dg-final { scan-tree-dump-not "__builtin_stack_save" "gimple" } } */
+/* { dg-final { scan-tree-dump-not "__builtin_alloca" "gimple" } } */
+/* { dg-final { scan-tree-dump-not "__builtin_stack_restore" "gimple" } } */
+
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_alloc \\(" 5 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(" 5 "gimple" } } */
+
+void
+one ()
+{
+ int result = 0, n = 3;
+ #pragma omp target map(tofrom: result) firstprivate(n)
+ {
+ int var = 5, var2[n];
+ #pragma omp allocate(var,var2) align(128) allocator(omp_low_lat_mem_alloc)
+/* { dg-final { scan-tree-dump-times "var\\.\[0-9\]+ = __builtin_GOMP_alloc \\(128, 4, 5\\);" 1 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "var2\\.\[0-9\]+ = __builtin_GOMP_alloc \\(128, D\\.\[0-9\]+, 5\\);" 1 "gimple" } } */
+
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(var\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(var2\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+
+ if ((intptr_t) &var % 128 != 0)
+ __builtin_abort ();
+ if ((intptr_t) var2 % 128 != 0)
+ __builtin_abort ();
+ if (var != 5)
+ __builtin_abort ();
+
+ #pragma omp parallel for
+ for (int i = 0; i < n; ++i)
+ var2[i] = (i+33);
+
+ #pragma omp loop reduction(+:result)
+ for (int i = 0; i < n; ++i)
+ result += var + var2[i];
+ }
+ if (result != (3*5 + 33 + 34 + 35))
+ __builtin_abort ();
+}
+
+void
+two ()
+{
+ struct st {
+ int a, b;
+ };
+ int scalar = 44, array[5] = {1,2,3,4,5};
+ struct st s = {.a=11, .b=56};
+ #pragma omp allocate(scalar, array, s)
+/* { dg-final { scan-tree-dump-times "scalar\\.\[0-9\]+ = __builtin_GOMP_alloc \\(4, 4, 0B\\);" 1 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "array\\.\[0-9\]+ = __builtin_GOMP_alloc \\(4, 20, 0B\\);" 1 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "s\\.\[0-9\]+ = __builtin_GOMP_alloc \\(4, 8, 0B\\);" 1 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(scalar\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(array\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(s\\.\[0-9\]+, 0B\\);" 1 "gimple" } } */
+
+ #pragma omp parallel firstprivate(scalar) firstprivate(array) firstprivate(s)
+ {
+ if (scalar != 44)
+ __builtin_abort ();
+ scalar = 33;
+ for (int i = 0; i < 5; ++i)
+ if (array[i] != i+1)
+ __builtin_abort ();
+ for (int i = 0; i < 5; ++i)
+ array[i] = 10*(i+1);
+ if (s.a != 11 || s.b != 56)
+ __builtin_abort ();
+ s.a = 74;
+ s.b = 674;
+ }
+ if (scalar != 44)
+ __builtin_abort ();
+ for (int i = 0; i < 5; ++i)
+ if (array[i] != i+1)
+ __builtin_abort ();
+ if (s.a != 11 || s.b != 56)
+ __builtin_abort ();
+
+ #pragma omp target defaultmap(firstprivate : scalar) defaultmap(none : aggregate) defaultmap(none : pointer)
+ {
+ if (scalar != 44)
+ __builtin_abort ();
+ scalar = 33;
+ }
+ if (scalar != 44)
+ __builtin_abort ();
+
+ #pragma omp target defaultmap(none : scalar) defaultmap(firstprivate : aggregate) defaultmap(none : pointer)
+ {
+ for (int i = 0; i < 5; ++i)
+ if (array[i] != i+1)
+ __builtin_abort ();
+ for (int i = 0; i < 5; ++i)
+ array[i] = 10*(i+1);
+ }
+ for (int i = 0; i < 5; ++i)
+ if (array[i] != i+1)
+ __builtin_abort ();
+ #pragma omp target defaultmap(none : scalar) defaultmap(firstprivate : aggregate) defaultmap(none : pointer)
+ {
+ if (s.a != 11 || s.b != 56)
+ __builtin_abort ();
+ s.a = 74;
+ s.b = 674;
+ }
+ if (s.a != 11 || s.b != 56)
+ __builtin_abort ();
+}
+
+int
+main ()
+{
+ one ();
+ two ();
+ return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c/allocate-6.c b/libgomp/testsuite/libgomp.c/allocate-6.c
new file mode 100644
index 00000000000..6d7278ce571
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c/allocate-6.c
@@ -0,0 +1,319 @@
+/* TODO: move to ../libgomp.c-c++-common once C++ is implemented. */
+/* NOTE: { target c } is unsupported with with the C compiler. */
+
+/* { dg-do run } */
+/* { dg-additional-options "-fdump-tree-omplower" } */
+
+/* For the 4 vars in omp_parallel, 4 in omp_target and 1 of 2 in no_alloc2_func. */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_alloc \\(" 9 "omplower" } } */
+/* { dg-final { scan-tree-dump-times "__builtin_GOMP_free \\(" 9 "omplower" } } */
+
+#include <omp.h>
+
+void
+check_int (int *x, int y)
+{
+ if (*x != y)
+ __builtin_abort ();
+}
+
+void
+check_ptr (int **x, int *y)
+{
+ if (*x != y)
+ __builtin_abort ();
+}
+
+
+int
+no_alloc_func ()
+{
+ /* There is no __builtin_GOMP_alloc / __builtin_GOMP_free as
+ allocator == omp_default_mem_alloc (known at compile time. */
+ int no_alloc;
+ #pragma omp allocate(no_alloc) allocator(omp_default_mem_alloc)
+ no_alloc = 7;
+ return no_alloc;
+}
+
+int
+no_alloc2_func()
+{
+ /* There is no __builtin_GOMP_alloc / __builtin_GOMP_free as
+ no_alloc2 is TREE_UNUSED. But there is for is_alloc2. */
+ int no_alloc2, is_alloc2;
+ #pragma omp allocate(no_alloc2, is_alloc2)
+ is_alloc2 = 7;
+ return is_alloc2;
+}
+
+
+void
+omp_parallel ()
+{
+ int n = 6;
+ int iii = 5, jjj[5], kkk[n];
+ int *ptr = (int *) 0x1234;
+ #pragma omp allocate(iii, jjj, kkk, ptr)
+
+ for (int i = 0; i < 5; i++)
+ jjj[i] = 3*i;
+ for (int i = 0; i < 6; i++)
+ kkk[i] = 7*i;
+
+ #pragma omp parallel default(none) firstprivate(iii, jjj, kkk, ptr) if(0)
+ {
+ if (iii != 5)
+ __builtin_abort();
+ iii = 7;
+ check_int (&iii, 7);
+ for (int i = 0; i < 5; i++)
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ for (int i = 0; i < 6; i++)
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ for (int i = 0; i < 5; i++)
+ jjj[i] = 4*i;
+ for (int i = 0; i < 6; i++)
+ kkk[i] = 8*i;
+ for (int i = 0; i < 5; i++)
+ check_int (&jjj[i], 4*i);
+ for (int i = 0; i < 6; i++)
+ check_int (&kkk[i], 8*i);
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ ptr = (int *) 0xabcd;
+ if (ptr != (int *) 0xabcd)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0xabcd);
+ }
+ if (iii != 5)
+ __builtin_abort ();
+ check_int (&iii, 5);
+ for (int i = 0; i < 5; i++)
+ {
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ check_int (&jjj[i], 3*i);
+ }
+ for (int i = 0; i < 6; i++)
+ {
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ check_int (&kkk[i], 7*i);
+ }
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0x1234);
+
+ #pragma omp parallel default(firstprivate) if(0)
+ {
+ if (iii != 5)
+ __builtin_abort();
+ iii = 7;
+ check_int (&iii, 7);
+ for (int i = 0; i < 5; i++)
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ for (int i = 0; i < 6; i++)
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ for (int i = 0; i < 5; i++)
+ jjj[i] = 4*i;
+ for (int i = 0; i < 6; i++)
+ kkk[i] = 8*i;
+ for (int i = 0; i < 5; i++)
+ check_int (&jjj[i], 4*i);
+ for (int i = 0; i < 6; i++)
+ check_int (&kkk[i], 8*i);
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ ptr = (int *) 0xabcd;
+ if (ptr != (int *) 0xabcd)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0xabcd);
+ }
+ if (iii != 5)
+ __builtin_abort ();
+ check_int (&iii, 5);
+ for (int i = 0; i < 5; i++)
+ {
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ check_int (&jjj[i], 3*i);
+ }
+ for (int i = 0; i < 6; i++)
+ {
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ check_int (&kkk[i], 7*i);
+ }
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0x1234);
+}
+
+
+
+void
+omp_target ()
+{
+ int n = 6;
+ int iii = 5, jjj[5], kkk[n];
+ int *ptr = (int *) 0x1234;
+ #pragma omp allocate(iii, jjj, kkk, ptr)
+
+ for (int i = 0; i < 5; i++)
+ jjj[i] = 3*i;
+ for (int i = 0; i < 6; i++)
+ kkk[i] = 7*i;
+
+ #pragma omp target defaultmap(none) firstprivate(iii, jjj, kkk, ptr)
+ {
+ if (iii != 5)
+ __builtin_abort();
+ iii = 7;
+ check_int (&iii, 7);
+ for (int i = 0; i < 5; i++)
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ for (int i = 0; i < 6; i++)
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ for (int i = 0; i < 5; i++)
+ jjj[i] = 4*i;
+ for (int i = 0; i < 6; i++)
+ kkk[i] = 8*i;
+ for (int i = 0; i < 5; i++)
+ check_int (&jjj[i], 4*i);
+ for (int i = 0; i < 6; i++)
+ check_int (&kkk[i], 8*i);
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ ptr = (int *) 0xabcd;
+ if (ptr != (int *) 0xabcd)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0xabcd);
+ }
+ if (iii != 5)
+ __builtin_abort ();
+ check_int (&iii, 5);
+ for (int i = 0; i < 5; i++)
+ {
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ check_int (&jjj[i], 3*i);
+ }
+ for (int i = 0; i < 6; i++)
+ {
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ check_int (&kkk[i], 7*i);
+ }
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0x1234);
+
+ #pragma omp target defaultmap(firstprivate)
+ {
+ if (iii != 5)
+ __builtin_abort();
+ iii = 7;
+ check_int (&iii, 7);
+ for (int i = 0; i < 5; i++)
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ for (int i = 0; i < 6; i++)
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ for (int i = 0; i < 5; i++)
+ jjj[i] = 4*i;
+ for (int i = 0; i < 6; i++)
+ kkk[i] = 8*i;
+ for (int i = 0; i < 5; i++)
+ check_int (&jjj[i], 4*i);
+ for (int i = 0; i < 6; i++)
+ check_int (&kkk[i], 8*i);
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ ptr = (int *) 0xabcd;
+ if (ptr != (int *) 0xabcd)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0xabcd);
+ }
+ if (iii != 5)
+ __builtin_abort ();
+ check_int (&iii, 5);
+ for (int i = 0; i < 5; i++)
+ {
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ check_int (&jjj[i], 3*i);
+ }
+ for (int i = 0; i < 6; i++)
+ {
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ check_int (&kkk[i], 7*i);
+ }
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0x1234);
+
+ #pragma omp target defaultmap(tofrom)
+ {
+ if (iii != 5)
+ __builtin_abort();
+ iii = 7;
+ check_int (&iii, 7);
+ for (int i = 0; i < 5; i++)
+ if (jjj[i] != 3*i)
+ __builtin_abort ();
+ for (int i = 0; i < 6; i++)
+ if (kkk[i] != 7*i)
+ __builtin_abort ();
+ for (int i = 0; i < 5; i++)
+ jjj[i] = 4*i;
+ for (int i = 0; i < 6; i++)
+ kkk[i] = 8*i;
+ for (int i = 0; i < 5; i++)
+ check_int (&jjj[i], 4*i);
+ for (int i = 0; i < 6; i++)
+ check_int (&kkk[i], 8*i);
+ if (ptr != (int *) 0x1234)
+ __builtin_abort ();
+ ptr = (int *) 0xabcd;
+ if (ptr != (int *) 0xabcd)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0xabcd);
+ }
+
+ if (iii != 7)
+ __builtin_abort ();
+ check_int (&iii, 7);
+ for (int i = 0; i < 5; i++)
+ {
+ if (jjj[i] != 4*i)
+ __builtin_abort ();
+ check_int (&jjj[i], 4*i);
+ }
+ for (int i = 0; i < 6; i++)
+ {
+ if (kkk[i] != 8*i)
+ __builtin_abort ();
+ check_int (&kkk[i], 8*i);
+ }
+ if (ptr != (int *) 0xabcd)
+ __builtin_abort ();
+ check_ptr (&ptr, (int *) 0xabcd);
+}
+
+
+int
+main ()
+{
+ omp_parallel ();
+ omp_target ();
+ return 0;
+}
^ permalink raw reply [flat|nested] only message in thread
only message in thread, other threads:[~2023-09-20 14:04 UTC | newest]
Thread overview: (only message) (download: mbox.gz / follow: Atom feed)
-- links below jump to the message on this page --
2023-09-20 14:04 [gcc r14-4176] OpenMP: Add ME support for 'omp allocate' stack variables Tobias Burnus
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).