https://gcc.gnu.org/g:9e9e22587f7954e079ac68e0e878f1feec7977f8
commit r17-2414-g9e9e22587f7954e079ac68e0e878f1feec7977f8 Author: Tobias Burnus <[email protected]> Date: Wed Jul 15 13:45:49 2026 +0200 OpenMP: libgomp.texi + allocate/uses_allocators janitorial updates [PR122748] * C only: Fix an ICE on invalid for the allocator clause (undeclared identifier); this is PR122748. * C/C++: Mark trait variables read to avoid bogus 'unused but set' warning with uses_allocators. * libgomp.texi: - Update implementation status for uses_allocators, claim that C23 is supported (minus bugs), and remove comment regarding a not-printed deprecation warning given that there is now one. - Memory Management Routines: Move 'omp_free' after the allocation routines for internal consistency and in line with the spec. Add '@c * omp_get_dyn_gprivate_...:: <fixme>/TR15' items to '@menu' for the new TR15 routines (actually TR14 but renamed). Hence, "grep '@c \*'" should yield all unimplemented or undocumented 'omp_' functions as of TR15. - nvptx: Fix an URL that was broken by Nvidia but the new one of commit r17-1942-g0e2c5d4b2c16c4 wasn't helpful. PR c/122748 gcc/c/ChangeLog: * c-parser.cc (c_parser_omp_clause_allocate): Avoid ICE when allocator is an error_mark_node. (c_parser_omp_clause_uses_allocators): Mark trait var as read to avoid 'unused but set' warning. gcc/cp/ChangeLog: * parser.cc (cp_parser_omp_clause_uses_allocators): Mark trait var as read to avoid 'unused but set' warning. libgomp/ChangeLog: * libgomp.texi (Implementation status): Mark two uses_allocators items and C23 support as 'Y'; remove outdated remark regarding a deprecation warning. (Memory Management Routines): Move omp_free after omp_realloc; add commented @menu entries for new TR15 routines. (nvptx): Update @uref to point to a better webpage. gcc/testsuite/ChangeLog: * c-c++-common/gomp/uses_allocators-1.c: New test. * c-c++-common/gomp/uses_allocators-2.c: New test. Diff: --- gcc/c/c-parser.cc | 16 ++- gcc/cp/parser.cc | 6 +- .../c-c++-common/gomp/uses_allocators-1.c | 74 ++++++++++++++ .../c-c++-common/gomp/uses_allocators-2.c | 23 +++++ libgomp/libgomp.texi | 108 +++++++++++---------- 5 files changed, 169 insertions(+), 58 deletions(-) diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc index 4eba29997273..9dc5ba1e3a13 100644 --- a/gcc/c/c-parser.cc +++ b/gcc/c/c-parser.cc @@ -19640,11 +19640,14 @@ c_parser_omp_clause_allocate (c_parser *parser, tree list) location_t expr_loc = c_parser_peek_token (parser)->location; c_expr expr = c_parser_expr_no_commas (parser, NULL); expr = convert_lvalue_to_rvalue (expr_loc, expr, false, true); - allocator = expr.value; - allocator = c_fully_fold (allocator, false, NULL); - orig_type = expr.original_type - ? expr.original_type : TREE_TYPE (allocator); - orig_type = TYPE_MAIN_VARIANT (orig_type); + if (expr.value != error_mark_node) + { + allocator = expr.value; + allocator = c_fully_fold (allocator, false, NULL); + orig_type = expr.original_type + ? expr.original_type : TREE_TYPE (allocator); + orig_type = TYPE_MAIN_VARIANT (orig_type); + } } if (allocator && (!INTEGRAL_TYPE_P (TREE_TYPE (allocator)) @@ -19790,6 +19793,7 @@ parse_next: { traits_var = expr.value; traits_var = c_fully_fold (traits_var, false, NULL); + mark_exp_read (traits_var); } else { @@ -19828,6 +19832,7 @@ parse_next: { traits_var = expr.value; traits_var = c_fully_fold (traits_var, false, NULL); + mark_exp_read (traits_var); } else { @@ -19879,6 +19884,7 @@ parse_next: legacy_traits = c_fully_fold (expr.value, false, NULL); if (legacy_traits == error_mark_node) goto end; + mark_exp_read (legacy_traits); gcc_rich_location richloc (make_location (alloc_loc, alloc_loc, close_loc)); if (nl == list) diff --git a/gcc/cp/parser.cc b/gcc/cp/parser.cc index 95e7544fff30..f455092b989e 100644 --- a/gcc/cp/parser.cc +++ b/gcc/cp/parser.cc @@ -44183,7 +44183,10 @@ parse_next: if (traits_var != NULL_TREE) dup_mod_tok = mod_tok; else - traits_var = t; + { + traits_var = t; + mark_exp_read (traits_var); + } } else { @@ -44268,6 +44271,7 @@ parse_next: legacy_traits = arg; if (legacy_traits == error_mark_node) goto end; + mark_exp_read (legacy_traits); gcc_rich_location richloc (make_location (tok->location, tok->location, close_loc)); if (nl == list) diff --git a/gcc/testsuite/c-c++-common/gomp/uses_allocators-1.c b/gcc/testsuite/c-c++-common/gomp/uses_allocators-1.c new file mode 100644 index 000000000000..8e8f2a861321 --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/uses_allocators-1.c @@ -0,0 +1,74 @@ +// { dg-do compile } +// { dg-additional-options "-Wall -Wno-deprecated-openmp" } + +// Check check that there is no 'unsed but set' warning for 'traits' + +// #include <omp.h> + +typedef __UINTPTR_TYPE__ omp_uintptr_t; + +typedef enum omp_alloctrait_key_t +{ + omp_atk_alignment = 2, +} omp_alloctrait_key_t; + +typedef enum omp_alloctrait_value_t +{ + omp_atv_default = (__UINTPTR_TYPE__) -1, +} omp_alloctrait_value_t; + +typedef struct omp_alloctrait_t +{ + omp_alloctrait_key_t key; + omp_uintptr_t value; +} omp_alloctrait_t; + +typedef enum omp_allocator_handle_t +{ + omp_null_allocator = 0, + __omp_allocator_handle_t_max__ = __UINTPTR_MAX__ +} omp_allocator_handle_t; + +typedef enum omp_memspace_handle_t +{ + omp_default_mem_space = 0, + omp_low_lat_mem_space = 4, + __omp_memspace_handle_t_max__ = __UINTPTR_MAX__ +} omp_memspace_handle_t; + + +extern void *omp_alloc (__SIZE_TYPE__, omp_allocator_handle_t); +extern void omp_free (void *, omp_allocator_handle_t); + +// ------------------------ + +int main() { + double x; // { dg-warning "variable 'x' set but not used" } + + const omp_alloctrait_t traits1[] = {{omp_atk_alignment, 1024}}; // { dg-bogus "variable 'traits1' set but not used" } + const omp_alloctrait_t traits2[] = {{omp_atk_alignment, 1024}}; // { dg-bogus "variable 'traits2' set but not used" } + const omp_alloctrait_t traits3[] = {{omp_atk_alignment, 1024}}; // { dg-bogus "variable 'traits3' set but not used" } + omp_allocator_handle_t alloc1 = omp_null_allocator; + omp_allocator_handle_t alloc2 = omp_null_allocator; + omp_allocator_handle_t alloc3 = omp_null_allocator; + + #pragma omp target \ + uses_allocators(traits(traits1), memspace(omp_low_lat_mem_space) : alloc1) \ + uses_allocators(memspace(omp_default_mem_space), traits(traits2) : alloc2) \ + uses_allocators(alloc3(traits3)) + { + __builtin_printf ("%p\n", (void*) alloc1); + __builtin_printf ("%p\n", (void*) alloc2); + __builtin_printf ("%p\n", (void*) alloc3); + int *ptr1 = (int*) omp_alloc(sizeof(int), alloc1); + int *ptr2 = (int*) omp_alloc(sizeof(int), alloc2); + int *ptr3 = (int*) omp_alloc(sizeof(int), alloc3); + x = 5; + *ptr1 = 5; + *ptr2 = 6; + *ptr3 = 7; + omp_free (ptr1, omp_null_allocator); + omp_free (ptr2, omp_null_allocator); + omp_free (ptr3, omp_null_allocator); + } +} diff --git a/gcc/testsuite/c-c++-common/gomp/uses_allocators-2.c b/gcc/testsuite/c-c++-common/gomp/uses_allocators-2.c new file mode 100644 index 000000000000..25added7fa2d --- /dev/null +++ b/gcc/testsuite/c-c++-common/gomp/uses_allocators-2.c @@ -0,0 +1,23 @@ +// { dg-do compile } + +// PR c/122748 +// Was previously ICEing + +void f() +{ + int x; +#pragma omp target parallel \ + uses_allocators(omp_thread_mem_alloc) \ + allocate(omp_thread_mem_alloc: x) private(x) + { } +} + +// C +// { dg-error "'omp_thread_mem_alloc' undeclared \\(first use in this function\\)" "" { target c } 10 } +// { dg-note "each undeclared identifier is reported only once for each function it appears in" "" { target c } 10 } +// { dg-error "allocate' clause must specify an allocator here" "" { target c } 11 } + +// C++ +// { dg-error "'omp_thread_mem_alloc' has not been declared" "" { target c++ } 10 } +// { dg-error "'omp_thread_mem_alloc' was not declared in this scope" "" { target c++ } 11 } +// { dg-error "'allocate' clause must specify an allocator here" "" { target c++ } 11 } diff --git a/libgomp/libgomp.texi b/libgomp/libgomp.texi index d090e2e86d63..687ae1a7cbca 100644 --- a/libgomp/libgomp.texi +++ b/libgomp/libgomp.texi @@ -416,14 +416,13 @@ to address of matching mapped list item per 5.1, Sect. 2.21.7.2 @tab N @tab allocatables/pointers @tab Y @tab @item Optional paired @code{end} directive with @code{dispatch} @tab Y @tab @item New @code{memspace} and @code{traits} modifiers for @code{uses_allocators} - @tab P @tab Only predefined allocators + @tab Y @tab @item Deprecation of traits array following the allocator_handle expression in @code{uses_allocators} @tab Y @tab @item New @code{otherwise} clause as alias for @code{default} on metadirectives @tab Y @tab @item Deprecation of @code{default} clause on metadirectives @tab Y - @tab Both @code{otherwise} and @code{default} are accepted - without diagnostics. + @tab @item Deprecation of delimited form of @code{declare target} @tab Y @tab @item Reproducible semantics changed for @code{order(concurrent)} @tab N @tab @item @code{allocate} and @code{firstprivate} clauses on @code{scope} @@ -463,7 +462,7 @@ to address of matching mapped list item per 5.1, Sect. 2.21.7.2 @tab N @tab @multitable @columnfractions .60 .10 .25 @item Features deprecated in versions 5.0, 5.1 and 5.2 were removed @tab N/A @tab Backward compatibility -@item Full support for C23 was added @tab P @tab +@item Full support for C23 was added @tab Y @tab @item Full support for C++23 was added @tab P @tab @item Full support for Fortran 2023 was added @tab P @tab @item @code{_ALL} suffix to the device-scope environment variables @@ -543,7 +542,7 @@ to address of matching mapped list item per 5.1, Sect. 2.21.7.2 @tab N @tab @item @code{access} allocator trait changes @tab N @tab @item New @code{partitioner} value to @code{partition} allocator trait @tab N @tab -@item Semicolon-separated list to @code{uses_allocators} @tab N @tab +@item Semicolon-separated list to @code{uses_allocators} @tab Y @tab @item New @code{need_device_addr} modifier to @code{adjust_args} clause @tab N @tab @item @code{interop} clause to @code{dispatch} @tab Y @tab @item Scope requirement changes for @code{declare_target} @tab N @tab @@ -3943,10 +3942,15 @@ They have C linkage and do not throw exceptions. * omp_get_default_allocator:: Get the default allocator * omp_alloc:: Memory allocation with an allocator * omp_aligned_alloc:: Memory allocation with an allocator and alignment -* omp_free:: Freeing memory allocated with OpenMP routines * omp_calloc:: Allocate nullified memory with an allocator * omp_aligned_calloc:: Allocate nullified aligned memory with an allocator * omp_realloc:: Reallocate memory allocated with OpenMP routines +* omp_free:: Freeing memory allocated with OpenMP routines +@c * omp_get_dyn_gprivate_ptr:: <fixme>/TR15 +@c * omp_get_dyn_gprivate_nofb_ptr:: <fixme>/TR15 +@c * omp_get_dyn_gprivate_size:: <fixme>/TR15 +@c * omp_get_dyn_gprivate_memspace:: <fixme>/TR15 +@c * omp_get_gprivate_limit:: <fixme>/TR15 @end menu @@ -4186,51 +4190,6 @@ Memory allocated by @code{omp_aligned_alloc} must be freed using -@node omp_free -@subsection @code{omp_free} -- Freeing memory allocated with OpenMP routines -@table @asis -@item @emph{Description}: -The @code{omp_free} routine deallocates memory previously allocated by an -OpenMP memory-management routine. The @var{ptr} argument must point to such -memory or be a null pointer; if it is a null pointer, no operation is -performed. If specified, the @var{allocator} argument must be either the -memory allocator that was used for the allocation or @code{omp_null_allocator}; -if it is @code{omp_null_allocator}, the implementation will determine the value -automatically. - -Calling @code{omp_free} invokes undefined behavior if the memory -was already deallocated or when the used allocator has already been destroyed. - -@item @emph{C}: -@multitable @columnfractions .20 .80 -@item @emph{Prototype}: @tab @code{void omp_free(void *ptr,} -@item @tab @code{ omp_allocator_handle_t allocator)} -@end multitable - -@item @emph{C++}: -@multitable @columnfractions .20 .80 -@item @emph{Prototype}: @tab @code{void omp_free(void *ptr,} -@item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator)} -@end multitable - -@item @emph{Fortran}: -@multitable @columnfractions .20 .80 -@item @emph{Interface}: @tab @code{subroutine omp_free(ptr, allocator) bind(C)} -@item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr} -@item @tab @code{type (c_ptr), value :: ptr} -@item @tab @code{integer (omp_allocator_handle_kind), value :: allocator} -@end multitable - -@item @emph{See also}: -@ref{omp_alloc}, @ref{omp_aligned_alloc}, @ref{omp_calloc}, -@ref{omp_aligned_calloc}, @ref{omp_realloc} - -@item @emph{Reference}: -@uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.7 -@end table - - - @node omp_calloc @subsection @code{omp_calloc} -- Allocate nullified memory with an allocator @table @asis @@ -4409,6 +4368,51 @@ was already deallocated or when the used allocator has already been destroyed. +@node omp_free +@subsection @code{omp_free} -- Freeing memory allocated with OpenMP routines +@table @asis +@item @emph{Description}: +The @code{omp_free} routine deallocates memory previously allocated by an +OpenMP memory-management routine. The @var{ptr} argument must point to such +memory or be a null pointer; if it is a null pointer, no operation is +performed. If specified, the @var{allocator} argument must be either the +memory allocator that was used for the allocation or @code{omp_null_allocator}; +if it is @code{omp_null_allocator}, the implementation will determine the value +automatically. + +Calling @code{omp_free} invokes undefined behavior if the memory +was already deallocated or when the used allocator has already been destroyed. + +@item @emph{C}: +@multitable @columnfractions .20 .80 +@item @emph{Prototype}: @tab @code{void omp_free(void *ptr,} +@item @tab @code{ omp_allocator_handle_t allocator)} +@end multitable + +@item @emph{C++}: +@multitable @columnfractions .20 .80 +@item @emph{Prototype}: @tab @code{void omp_free(void *ptr,} +@item @tab @code{ omp_allocator_handle_t allocator=omp_null_allocator)} +@end multitable + +@item @emph{Fortran}: +@multitable @columnfractions .20 .80 +@item @emph{Interface}: @tab @code{subroutine omp_free(ptr, allocator) bind(C)} +@item @tab @code{use, intrinsic :: iso_c_binding, only : c_ptr} +@item @tab @code{type (c_ptr), value :: ptr} +@item @tab @code{integer (omp_allocator_handle_kind), value :: allocator} +@end multitable + +@item @emph{See also}: +@ref{omp_alloc}, @ref{omp_aligned_alloc}, @ref{omp_calloc}, +@ref{omp_aligned_calloc}, @ref{omp_realloc} + +@item @emph{Reference}: +@uref{https://www.openmp.org, OpenMP specification v5.0}, Section 3.7.7 +@end table + + + @node Tool Control Routine @section Tool Control Routine @@ -7888,7 +7892,7 @@ The implementation remark: @item OpenMP code that has a @code{requires} directive with @code{self_maps} or @code{unified_shared_memory} runs on nvptx devices if and only if all of those support the @code{pageableMemoryAccess} property;@footnote{ - @uref{https://docs.nvidia.com/cuda/cuda-programming-guide/}} + @uref{https://docs.nvidia.com/cuda/cuda-programming-guide/04-special-topics/unified-memory.html}} otherwise, all nvptx device are removed from the list of available devices (``host fallback''). @item The default per-warp stack size is 128 kiB; see also @code{-msoft-stack}
