OpenMP: Handle more cases in user/condition selector
Tobias had noted that the C front end was not treating C23 constexprs
as constant in the user/condition selector property, which led to
missed opportunities to resolve metadirectives at parse time.
Additionally neither C nor C++ was permitting the expression to have
pointer or floating-point type -- the former being a common idiom in
other C/C++ conditional expressions. By using the existing front-end
hooks for the implicit conversion to bool in conditional expressions,
we also get free support for using a C++ class object that has a bool
conversion operator in the user/condition selector.
gcc/c/ChangeLog
* c-parser.cc (c_parser_omp_context_selector): Call
convert_lvalue_to_rvalue and c_objc_common_truthvalue_conversion
on the expression for OMP_TRAIT_PROPERTY_BOOL_EXPR.
gcc/cp/ChangeLog
* cp-tree.h (maybe_convert_cond): Declare.
* parser.cc (cp_parser_omp_context_selector): Call
maybe_convert_cond and fold_build_cleanup_point_expr on the
expression for OMP_TRAIT_PROPERTY_BOOL_EXPR.
* pt.cc (tsubst_omp_context_selector): Likewise.
* semantics.cc (maybe_convert_cond): Remove static declaration.
Thomas Schwinge [Mon, 26 May 2025 11:31:54 +0000 (13:31 +0200)]
Avoid SIGSEGV in nvptx 'mkoffload' for voluminous PTX code
In commit 50be486dff4ea2676ed022e9524ef190b92ae2b1
"nvptx: libgomp+mkoffload.cc: Prepare for reverse offload fn lookup", some
additional tracking of the PTX code was added, and this assumes that
potentially every single character of PTX code needs to be tracked as a new
chunk of PTX code. That's problematic if we're dealing with voluminous PTX
code (for example, non-trivial C++ code), and the 'file_idx' 'alloca'tion then
causes stack overflow. For example:
FAIL: libgomp.c++/target-std__valarray-1.C (test for excess errors)
UNRESOLVED: libgomp.c++/target-std__valarray-1.C compilation failed to produce executable
lto-wrapper: fatal error: [...]/build-gcc/gcc//accel/nvptx-none/mkoffload terminated with signal 11 [Segmentation fault], core dumped
gcc/
* config/nvptx/mkoffload.cc (process): Use an 'auto_vec' for
'file_idx'.
Jerry DeLisle [Sat, 31 May 2025 15:57:22 +0000 (08:57 -0700)]
Fortran: Fix handling of parsed format strings.
Previously parsed strings with errors were being cached such
that subsequent use of the format string were not being
checked for errors.
PR libfortran/119856
libgfortran/ChangeLog:
* io/format.c (parse_format_list): Set the fmt->error
message for missing comma.
(parse_format): Do not cache the parsed format string
if a previous error ocurred.
Jerry DeLisle [Wed, 28 May 2025 14:56:12 +0000 (07:56 -0700)]
Fortran: Adjust handling of optional comma in FORMAT.
This change adjusts the error messages for optional commas
in format strings to give a warning at compile time unless
-std=legacy is used. This is more consistant with the
runtime library. A missing comma separator should not be
encouraged as it is non-standard fortran.
PR fortran/119856
gcc/fortran/ChangeLog:
* io.cc: Set missing comma error checks to STD_STD_LEGACY.
Patrick Palka [Thu, 29 May 2025 14:12:23 +0000 (10:12 -0400)]
libstdc++: Compare keys and values separately in flat_map::operator==
Instead of effectively doing a zipped comparison of the keys and values,
compare them separately to leverage the underlying containers' optimized
equality implementations.
libstdc++-v3/ChangeLog:
* include/std/flat_map (_Flat_map_impl::operator==): Compare
keys and values separately.
Patrick Palka [Thu, 29 May 2025 14:11:57 +0000 (10:11 -0400)]
libstdc++: Fix tuple/pair confusion with std::erase_if(flat_map) [PR120465]
std::erase_if for flat_map/multimap is implemented via ranges::erase_if
over a zip_view of the keys and values, the value_type of which is a
tuple, but the given predicate needs to be called with a pair (flat_map's
value_type). So use a projection to convert the tuple into a suitable
pair.
PR libstdc++/120465
libstdc++-v3/ChangeLog:
* include/std/flat_map (_Flat_map_impl::_M_erase_if): Use a
projection with ranges::remove_if to pass a pair instead of
a tuple to the predicate.
* testsuite/23_containers/flat_map/1.cc (test07): Strengthen
to expect the argument passed to the predicate is a pair.
* testsuite/23_containers/flat_multimap/1.cc (test07): Likewise.
Co-authored-by: Jonathan Wakely <jwakely@redhat.com> Reviewed-by: Tomasz KamiĆski <tkaminsk@redhat.com> Reviewed-by: Jonathan Wakely <jwakely@redhat.com>
(cherry picked from commit 589b27ec5769410e036df57645ff1eb7c765f692)
Jerry DeLisle [Tue, 20 May 2025 02:41:16 +0000 (19:41 -0700)]
Fortran: Fix c_associated argument checks.
PR fortran/120049
gcc/fortran/ChangeLog:
* check.cc (gfc_check_c_associated): Use new helper functions.
Only call check_c_ptr_1 if optional c_ptr_2 tests succeed.
(check_c_ptr_1): Handle only c_ptr_1 checks.
(check_c_ptr_2): Expand checks for c_ptr_2 and handle cases
where there is no derived pointer in the gfc_expr and check
the inmod_sym_id only if it exists. Rephrase error message.
* misc.cc (gfc_typename): Handle the case for BT_VOID rather
than throw an internal error.
gcc/testsuite/ChangeLog:
* gfortran.dg/pr120049_a.f90: Update test directives.
* gfortran.dg/pr120049_b.f90: Update test directives
* gfortran.dg/pr120049_2.f90: New test.
* gfortran.dg/c_f_pointer_tests_6.f90: Adjust dg-error
directive.
Thomas Koenig [Fri, 30 May 2025 11:31:58 +0000 (13:31 +0200)]
Type mismatch for passed external function
This obvious and simple patch fixes a 15/16 regression where the
typespec of a global function was in the RESULT clause and not
in the symbol itself.
gcc/fortran/ChangeLog:
PR fortran/120355
* interface.cc (compare_parameter): If the global function has a
result clause, take typespec from there for the comparison against
the dummy argument.
gcc/testsuite/ChangeLog:
PR fortran/120355
* gfortran.dg/interface_62.f90: New test.
Sandra Loosemore [Mon, 26 May 2025 19:21:48 +0000 (19:21 +0000)]
OpenMP: Fix ICE and other issues in C/C++ metadirective error recovery.
The new testcase included in this patch used to ICE in gcc after
diagnosing the first error, and in g++ it only diagnosed the error in
the first metadirective, ignoring the second one. The solution is to
make error recovery in the C front end more like that in the C++ front
end, and remove the code in both front ends that previously tried to
skip all the way over the following statement (instead of just to the
end of the metadirective pragma) after an error.
gcc/c/ChangeLog
* c-parser.cc (c_parser_skip_to_closing_brace): New, copied from
the equivalent function in the C++ front end.
(c_parser_skip_to_end_of_block_or_statement): Pass false to
the error flag.
(c_parser_omp_context_selector): Immediately return error_mark_node
after giving an error that the integer trait property is invalid,
similarly to C++ front end.
(c_parser_omp_context_selector_specification): Likewise handle
error return from c_parser_omp_context_selector similarly to C++.
(c_parser_omp_metadirective): Do not call
c_parser_skip_to_end_of_block_or_statement after an error.
gcc/cp/ChangeLog
* parser.cc (cp_parser_omp_metadirective): Do not call
cp_parser_skip_to_end_of_block_or_statement after an error.
gcc/testsuite/ChangeLog
* c-c++-common/gomp/declare-variant-2.c: Adjust patterns now that
C and C++ now behave similarly.
* c-c++-common/gomp/metadirective-error-recovery.c: New.
Waffl3x [Mon, 26 May 2025 08:38:27 +0000 (02:38 -0600)]
Add 'libgomp.c++/target-flex-[...].C' test cases
libgomp/ChangeLog:
* testsuite/libgomp.c++/target-flex-10.C: New test.
* testsuite/libgomp.c++/target-flex-100.C: New test.
* testsuite/libgomp.c++/target-flex-101.C: New test.
* testsuite/libgomp.c++/target-flex-11.C: New test.
* testsuite/libgomp.c++/target-flex-12.C: New test.
* testsuite/libgomp.c++/target-flex-2000.C: New test.
* testsuite/libgomp.c++/target-flex-2001.C: New test.
* testsuite/libgomp.c++/target-flex-2002.C: New test.
* testsuite/libgomp.c++/target-flex-2003.C: New test.
* testsuite/libgomp.c++/target-flex-30.C: New test.
* testsuite/libgomp.c++/target-flex-300.C: New test.
* testsuite/libgomp.c++/target-flex-31.C: New test.
* testsuite/libgomp.c++/target-flex-32.C: New test.
* testsuite/libgomp.c++/target-flex-33.C: New test.
* testsuite/libgomp.c++/target-flex-41.C: New test.
* testsuite/libgomp.c++/target-flex-60.C: New test.
* testsuite/libgomp.c++/target-flex-61.C: New test.
* testsuite/libgomp.c++/target-flex-62.C: New test.
* testsuite/libgomp.c++/target-flex-70.C: New test.
* testsuite/libgomp.c++/target-flex-80.C: New test.
* testsuite/libgomp.c++/target-flex-81.C: New test.
* testsuite/libgomp.c++/target-flex-90.C: New test.
* testsuite/libgomp.c++/target-flex-common.h: New test.
Tobias Burnus [Wed, 28 May 2025 13:14:14 +0000 (15:14 +0200)]
libgomp.fortran/metadirective-1.f90: Expect 'error:' for nvptx compile [PR118694]
This should have been part of commit r16-838-gb3d07ec7ac2ccd or r16-883-g5d6ed6d604ff94 - all showing the same issue:
'!$omp target' followed by a metadirective with 'teams'; if
the metadirective cannot be early resolved, a diagnostic
error is shown about using directives between 'target' and
'teams'.
While the message is misleading, the problem is that the
host invokes 'target' differently when 'teams' is present;
in this case, host fallback + amdgcn offload require the
no-teams case, nvptx offload the teams case such that it
only can be resolved at runtime.
Mark the error as 'dg-bogus + xfail' to silence the FAIL,
when nvptx offloading is compiled for. (If not, the
metadirective can be resolved early during compilation.)
libgomp/ChangeLog:
PR middle-end/118694
* testsuite/libgomp.fortran/metadirective-1.f90: xfail when
compiling (also) for nvptx offloading as an error is then expected.
Eric Botcazou [Tue, 27 May 2025 17:42:17 +0000 (19:42 +0200)]
Fix IPA-SRA issue with reverse SSO on specific pattern
IPA-SRA generally works fine in the presence of reverse Scalar_Storage_Order
by propagating the relevant flag onto the newly generated MEM_REFs. However
we have been recently faced with a specific Ada pattern that it does not
handle correctly: the 'Valid attribute applied to a floating-point component
of an aggregate type with reverse Scalar_Storage_Order.
The attribute is implemented by a call to a specific routine of the runtime
that expects a pointer to the object so, in the case of a component with
reverse SSO, the compiler first loads it from the aggregate to get back the
native storage order, but it does the load using an array of bytes instead
of the floating-point type to prevent the FPU from fiddling with the value,
which yields in the .original dump file:
Of course that's a bit convoluted, but it does not seem that another method
would be simpler or even work, and using VIEW_CONVERT_EXPR to toggle the SSO
is supposed to be supported in any case (unlike aliasing or type punning).
The attached patch makes it work. While the call to storage_order_barrier_p
from IPA-SRA is quite natural (the regular SRA has it too), the tweak to the
predicate itself is needed to handle the scalar->aggregate conversion, which
is admittedly awkward but again without clear alternative.
gcc/
* ipa-sra.cc (scan_expr_access): Also disqualify storage order
barriers from splitting.
* tree.h (storage_order_barrier_p): Also return false if the
operand of the VIEW_CONVERT_EXPR has reverse storage order.
gcc/testsuite/
* gnat.dg/sso19.adb: New test.
* gnat.dg/sso19_pkg.ads, gnat.dg/sso19_pkg.adb: New helper.
Jonathan Wakely [Wed, 21 May 2025 14:29:02 +0000 (15:29 +0100)]
libstdc++: Fix vector(from_range_t, R&&) for exceptions [PR120367]
Because this constructor delegates to vector(a) the object has been
fully constructed and the destructor will run if an exception happens.
That means we need to set _M_finish == _M_start so that the destructor
doesn't try to destroy any elements.
libstdc++-v3/ChangeLog:
PR libstdc++/120367
* include/bits/stl_vector.h (_M_range_initialize): Initialize
_M_impl._M_finish.
* testsuite/23_containers/vector/cons/from_range.cc: Check with
a type that throws on construction.
exceptions during construction.
Tobias Burnus [Mon, 26 May 2025 17:50:40 +0000 (19:50 +0200)]
c-c++-common/gomp/{attrs-,}metadirective-3.c: Fix expected result [PR118694]
With compilation for nvptx enabled, two issues showed up:
(a) "error: 'target' construct with nested 'teams' construct contains
directives outside of the 'teams' construct"
See PR comment 9 why this is difficult to fix.
Solution: Add dg-bogus and accept/expect the error for 'target offload_nvptx'.
(b) The assumptions about the dump for 'target offload_nvptx' were wrong
as the metadirective was already expanded to a OMP_NEXT_VARIANT
construct such that no 'omp metadirective' was left in either case.
Solution: Check that no 'omp metadirective' is left; additionally, expect
either OMP_NEXT_VARIANT (when offload_nvptx is available) or no 'teams'
directive at all (if not).
gcc/testsuite/ChangeLog:
PR middle-end/118694
* c-c++-common/gomp/attrs-metadirective-3.c: Change to never
expect 'omp metadirective' in the dump. If !offload_nvptx, check
that no 'teams' shows up in the dump; for offload_nvptx, expect
OMP_NEXT_VARIANT and an error about directive between 'target'
and 'teams'.
* c-c++-common/gomp/metadirective-3.c: Likewise.
Tobias Burnus [Fri, 23 May 2025 09:30:48 +0000 (11:30 +0200)]
libgomp.c-c++-common/metadirective-1.c: Expect 'error:' for nvptx compile [PR118694]
OpenMP's 'target teams' is strictly coupled with 'teams'; if the latter
exists, the kernel is launched in directly with multiple teams. Thus,
the host has to know whether the teams construct exists or not. For
#pragma omp target
#pragma omp metadirective when (device={arch("nvptx")}: teams loop)
it is simple when 'nvptx' offloading is not supported, otherwise it depends
on the default device at runtime as the user code asks for a single team for
host fallback and gcn offload and multiple for nvptx offload.
In any case, this commit ensures that no FAIL is printed, whatever a
future solution might look like. Instead of a dg-bogus combined with an
'xfail offload_target_nvptx', one an also argue that a dg-error for
'target offload_target_nvptx' would be more appropriate.
libgomp/ChangeLog:
PR middle-end/118694
* testsuite/libgomp.c-c++-common/metadirective-1.c: xfail when
compiling (also) for nvptx offloading as an error is then expected.
Nathaniel Shead [Thu, 22 May 2025 12:16:22 +0000 (22:16 +1000)]
c++/modules: Fix merge of TLS init functions [PR120363]
The PR notes that we missed setting DECL_CONTEXT on the TLS init
function; we missed this initially because this function is not created
in header units, only named modules.
I also noticed that 'DECL_CONTEXT (fn) = DECL_CONTEXT (var)' was
incorrect: for class members, this ends up having the modules merging
machinery treat the decl as a member function, which breaks when
attempting to dedup against an existing completed class type. Instead
we can just use the global_namespace as the context, because the name of
the function is already mangled appropriately so that we'll match the
correct duplicates.
PR c++/120363
gcc/cp/ChangeLog:
* decl2.cc (get_tls_init_fn): Set context as global_namespace.
(get_tls_wrapper_fn): Likewise.
gcc/testsuite/ChangeLog:
* g++.dg/modules/pr113292_a.H: Move to...
* g++.dg/modules/tls-1_a.H: ...here.
* g++.dg/modules/pr113292_b.C: Move to...
* g++.dg/modules/tls-1_b.C: ...here.
* g++.dg/modules/pr113292_c.C: Move to...
* g++.dg/modules/tls-1_c.C: ...here.
* g++.dg/modules/tls-2_a.C: New test.
* g++.dg/modules/tls-2_b.C: New test.
* g++.dg/modules/tls-2_c.C: New test.
* g++.dg/modules/tls-3.h: New test.
* g++.dg/modules/tls-3_a.H: New test.
* g++.dg/modules/tls-3_b.C: New test.
Signed-off-by: Nathaniel Shead <nathanieloshead@gmail.com> Reviewed-by: Jason Merrill <jason@redhat.com>
(cherry picked from commit 66e9a4f3083356b064cc64651edad466a56f762b)
Nathaniel Shead [Fri, 23 May 2025 14:51:49 +0000 (00:51 +1000)]
c++/modules: Fix stream-in of member using-decls [PR120414]
When streaming in a reference to a data member, we have an oversight
where we did not consider USING_DECLs, despite otherwise handling them
here the same as fields. This patch corrects that mistake.
PR c++/120414
gcc/cp/ChangeLog:
* module.cc (trees_in::tree_node): Allow reading a USING_DECL
when streaming tt_data_member.
gcc/testsuite/ChangeLog:
* g++.dg/modules/using-31_a.C: New test.
* g++.dg/modules/using-31_b.C: New test.
Signed-off-by: Nathaniel Shead <nathanieloshead@gmail.com> Reviewed-by: Jason Merrill <jason@redhat.com>
(cherry picked from commit 43dddeef7a870ce4db7407f73660504b67a0a919)
Harald Anlauf [Thu, 15 May 2025 19:07:07 +0000 (21:07 +0200)]
Fortran: default-initialization and functions returning derived type [PR85750]
Functions with non-pointer, non-allocatable result and of derived type did
not always get initialized although the type had default-initialization,
and a derived type component had the allocatable or pointer attribute.
Rearrange the logic when to apply default-initialization.
PR fortran/85750
gcc/fortran/ChangeLog:
* resolve.cc (resolve_symbol): Reorder conditions when to apply
default-initializers.
/* Handle empty records as per the x86-64 psABI. */
TYPE_EMPTY_P (type) = targetm.calls.empty_record_p (type);
(Indeed x86_64 is still the only target to define 'TARGET_EMPTY_RECORD_P',
calling 'gcc/tree.cc-default_is_empty_record'.)
And so it happens that for an empty struct used in code offloaded from x86_64
host (but not powerpc64le host, for example), we get to see 'TYPE_EMPTY_P' in
offloading compilation (where the offload targets (currently?) don't use it
themselves, and therefore aren't prepared to handle it).
For nvptx offloading compilation, this causes wrong code generation:
'ptxas [...] error : Call has wrong number of parameters', as nvptx code
generation for function definition doesn't pay attention to this flag (say, in
'gcc/config/nvptx/nvptx.cc:pass_in_memory', or whereever else would be
appropriate to handle that), but the generic code 'gcc/calls.cc:expand_call'
via 'gcc/function.cc:aggregate_value_p' does pay attention to it, and we thus
get mismatching function definition vs. function call.
This issue apparently isn't a problem for GCN offloading, but I don't know if
that's by design or by accident.
Richard Biener:
> It looks like TYPE_EMPTY_P is only used during RTL expansion for ABI
> purposes, so computing it during layout_type is premature as shown here.
>
> I would suggest to simply re-compute it at offload stream-in time.
(For avoidance of doubt, the additions to 'gcc.target/nvptx/abi-struct-arg.c',
'gcc.target/nvptx/abi-struct-ret.c' are not dependent on the offload streaming
code changes, but are just to mirror the changes to
'libgomp.oacc-c-c++-common/abi-struct-1.c'.)
Tobias Burnus [Thu, 22 May 2025 15:06:38 +0000 (17:06 +0200)]
git_repository.py: Fix handling of --last-commit with merges
contrib/ChangeLog:
* gcc-changelog/git_update_version.py (update_current_branch): Fix
handling of merges.
* gcc-changelog/git_repository.py (parse_git_revisions): Skip over
merge commits for merges from exclude branch.
Martin Jambor [Wed, 14 May 2025 10:08:24 +0000 (12:08 +0200)]
tree-sra: Do not create stores into const aggregates (PR111873)
This patch fixes (hopefully the) one remaining place where gimple SRA
was still creating a load into const aggregates. It occurs when there
is a replacement for a load but that replacement is not type
compatible - typically because it is a single field structure.
I have used testcases from duplicates because the original test-case
no longer reproduces for me.
gcc/ChangeLog:
2025-05-13 Martin Jambor <mjambor@suse.cz>
PR tree-optimization/111873
* tree-sra.cc (sra_modify_expr): When processing a load which has
a type-incompatible replacement, do not store the contents of the
replacement into the original aggregate when that aggregate is
const.
gcc/testsuite/ChangeLog:
2025-05-13 Martin Jambor <mjambor@suse.cz>
* gcc.dg/ipa/pr120044-1.c: New test.
* gcc.dg/ipa/pr120044-2.c: Likewise.
* gcc.dg/tree-ssa/pr114864.c: Likewise.
Martin Jambor [Fri, 16 May 2025 15:13:51 +0000 (17:13 +0200)]
ipa: Dump cgraph_node UID instead of order into ipa-clones dump file
Since starting from GCC 15 the order is not unique for any
symtab_nodes but m_uid is, I believe we ought to dump the latter in
the ipa-clones dump, if only so that people can reliably match entries
about new clones to those about removed nodes (if any).
This patch also contains a fixes to a few other places where we have
so far dumped order to our ordinary dumps and which have been
identified by Michal Jires.
gcc/ChangeLog:
2025-05-16 Martin Jambor <mjambor@suse.cz>
* cgraph.h (symtab_node): Make member function get_uid const.
* cgraphclones.cc (dump_callgraph_transformation): Dump m_uid of the
call graph nodes instead of order.
* cgraph.cc (cgraph_node::remove): Likewise.
* ipa-cp.cc (ipcp_lattice<valtype>::print): Likewise.
* ipa-sra.cc (ipa_sra_summarize_function): Likewise.
* symtab.cc (symtab_node::dump_base): Likewise.
Jonathan Wakely [Tue, 20 May 2025 09:53:41 +0000 (10:53 +0100)]
libstdc++: Fix incorrect links to archived SGI STL docs
In r8-7777-g25949ee33201f2 I updated some URLs to point to copies of the
SGI STL docs in the Wayback Machine, because the original pags were no
longer hosted on sgi.com. However, I incorrectly assumed that if one
archived page was at https://web.archive.org/web/20171225062613/... then
all the other pages would be too. Apparently that's not how the Wayback
Machine works, and each page is archived on a different date. That meant
that some of our links were redirecting to archived copies of the
announcement that the SGI STL docs have gone away.
This fixes each URL to refer to a correctly archived copy of the
original docs.
Nathaniel Shead [Sat, 17 May 2025 13:51:07 +0000 (23:51 +1000)]
c++/modules: Fix ICE on merge of instantiation with partial spec [PR120013]
When we import a pending instantiation that matches an existing partial
specialisation, we don't find the slot in the entity map because for
partial specialisations we register the TEMPLATE_DECL but for normal
implicit instantiations we instead register the inner TYPE_DECL.
Because the DECL_MODULE_ENTITY_P flag is set we correctly realise that
it is in the entity map, but ICE when attempting to use that slot in
partition handling.
This patch fixes the issue by detecting this case and instead looking
for the slot for the TEMPLATE_DECL. It doesn't matter that we never add
a slot for the inner decl because we're about to discard it anyway.
PR c++/120013
gcc/cp/ChangeLog:
* module.cc (trees_in::install_entity): Handle re-registering
the inner TYPE_DECL of a partial specialisation.
gcc/testsuite/ChangeLog:
* g++.dg/modules/partial-8.h: New test.
* g++.dg/modules/partial-8_a.C: New test.
* g++.dg/modules/partial-8_b.C: New test.
* g++.dg/modules/partial-8_c.C: New test.
* g++.dg/modules/partial-8_d.C: New test.
Signed-off-by: Nathaniel Shead <nathanieloshead@gmail.com> Reviewed-by: Jason Merrill <jason@redhat.com>
(cherry picked from commit b0de7297f2b5670386472229ab795a577c288ecf)
Nathaniel Shead [Mon, 19 May 2025 13:17:16 +0000 (23:17 +1000)]
c++/modules: Always mark tinfo vars as TREE_ADDRESSABLE [PR120350]
We need to mark type info decls as addressable if we take them by
reference; this is done by walking the declaration during parsing and
marking the decl as needed.
However, with modules we don't stream tinfo decls directly; rather we
stream just their name and type and reconstruct them in the importer
directly. This means that any addressable flags are not propagated, and
we error because TREE_ADDRESSABLE is not set despite taking its address.
But tinfo decls should always have TREE_ADDRESSABLE set, as any attempt
to use the tinfo decl will go through build_address anyway. So this
patch fixes the issue by eagerly marking the constructed decl as
TREE_ADDRESSABLE so that modules gets this flag correctly set as well.
PR c++/120350
gcc/cp/ChangeLog:
* rtti.cc (get_tinfo_decl_direct): Mark TREE_ADDRESSABLE.
gcc/testsuite/ChangeLog:
* g++.dg/modules/tinfo-3_a.H: New test.
* g++.dg/modules/tinfo-3_b.C: New test.
Signed-off-by: Nathaniel Shead <nathanieloshead@gmail.com> Reviewed-by: Jason Merrill <jason@redhat.com>
(cherry picked from commit 9a6e5a437f0416627ee516f6ef5929cb30c5e498)
Jonathan Wakely [Fri, 16 May 2025 10:54:46 +0000 (11:54 +0100)]
libstdc++: Fix some Clang -Wsystem-headers warnings in <ranges>
libstdc++-v3/ChangeLog:
* include/std/ranges (_ZipTransform::operator()): Remove name of
unused parameter.
(chunk_view::_Iterator, stride_view::_Iterator): Likewise.
(join_with_view): Declare _Iterator and _Sentinel as class
instead of struct.
(repeat_view): Declare _Iterator as class instead of struct.
Jonathan Wakely [Thu, 15 May 2025 18:32:01 +0000 (19:32 +0100)]
libstdc++: Fix std::format of chrono::local_days with {} [PR120293]
Formatting of chrono::local_days with an empty chrono-specs should be
equivalent to inserting it into an ostream, which should use the
overload for inserting chrono::sys_days into an ostream. The
implementation of empty chrono-specs in _M_format_to_ostream takes some
short cuts, and that wasn't being done correctly for chrono::local_days.
libstdc++-v3/ChangeLog:
PR libstdc++/120293
* include/bits/chrono_io.h (_M_format_to_ostream): Add special
case for local_time convertible to local_days.
* testsuite/std/time/clock/local/io.cc: Check formatting of
chrono::local_days.
Jonathan Wakely [Wed, 30 Apr 2025 16:31:01 +0000 (17:31 +0100)]
libstdc++: Fix dangling pointer in fs::path::operator+=(*this) [PR120029]
When concatenating a path we reallocate the left operand's storage to
make room for the new components being added. When the two operands are
the same object, or the right operand is one of the components of the
left operand, the reallocation invalidates the pointers that refer
into the right operand's storage.
The solution in this commit is to detect these aliasing cases and just
do the concatenation in terms of the contained string, as that code
already handles the case where the string aliases the path. The standard
specifies the concatenation in terms of the native() string, so all this
change does is disable the optimized implementation of concatenation for
path objects which attempts to avoid re-parsing the path from the
concatenated string.
The potential loss of performance for this case isn't likely to be an
issue, because concatenating a path with itself (or one of its existing
components) probably isn't a common use case.
The Filesystem TS implementation doesn't have the optimized form of
concatenation and always does it in terms of the native string and
reparsing the whole thing, so doesn't have this bug. A test is added to
confirm that anyway (that test has some slightly different results due
to different behaviour for trailing slashes and implicit "." filenames
in the TS spec).
libstdc++-v3/ChangeLog:
PR libstdc++/120029
* src/c++17/fs_path.cc (path::operator+=(const path&)): Handle
parameters that alias the path or one of its components.
* testsuite/27_io/filesystem/path/concat/120029.cc: New test.
* testsuite/experimental/filesystem/path/concat/120029.cc: New
test.
Jonathan Wakely [Thu, 15 May 2025 10:01:05 +0000 (11:01 +0100)]
libstdc++: Fix std::format_kind primary template for Clang [PR120190]
Although Clang trunk has been adjusted to handle our std::format_kind
definition (because they need to be able to compile the GCC 15.1.0
release), it's probably better to not rely on something that they might
start diagnosing again in future.
Define the primary template in terms of an immediately invoked function
expression, so that we can put a static_assert(false) in the body.
libstdc++-v3/ChangeLog:
PR libstdc++/120190
* include/std/format (format_kind): Adjust primary template to
not depend on itself.
* testsuite/std/format/ranges/format_kind_neg.cc: Adjust
expected errors. Check more invalid specializations.
Reviewed-by: Tomasz KamiĆski <tkaminsk@redhat.com> Reviewed-by: Daniel KrĂŒgler <daniel.kruegler@gmail.com>
(cherry picked from commit c65725eccbabf3b9b5965f27fff2d3b9f6c75930)
Tobias Burnus [Thu, 15 May 2025 07:15:21 +0000 (09:15 +0200)]
OpenMP/Fortran: Fix allocatable-component mapping of derived-type array comps
The check whether the location expression in map clause has allocatable
components was failing for some derived-type array expressions such as
map(var%tiles(1))
as the compiler produced
_4 = var.tiles;
MEMREF(_4, _5);
This commit now also handles this case.
gcc/fortran/ChangeLog:
* trans-openmp.cc (gfc_omp_deep_mapping_do): Handle SSA_NAME if
a def_stmt is available.
libgomp/ChangeLog:
* testsuite/libgomp.fortran/alloc-comp-4.f90: New test.
Tobias Burnus [Wed, 14 May 2025 18:06:49 +0000 (20:06 +0200)]
OpenMP: Fix mapping of zero-sized arrays with non-literal size: map(var[:n]), n = 0
For map(ptr[:0]), the used map kind is GOMP_MAP_ATTACH_ZERO_LENGTH_ARRAY_SECTION
and it is permitted that 'ptr' does not exist. 'ptr' is set to the device
pointee if it exists or to the host value otherwise.
For map(ptr[:3]), the variable is first mapped and then ptr is updated to point
to the just-mapped device data; the attachment uses GOMP_MAP_ATTACH.
For map(ptr[:n]), generates always a GOMP_MAP_ATTACH, but when n == 0, it
was failing with:
"pointer target not mapped for attach"
The solution is not to fail but first to check whether it was mapped before.
It turned out that for the mapping part, GCC adds a run-time check whether
n == 0 - and uses GOMP_MAP_ZERO_LEN_ARRAY_SECTION for the mapping.
Thus, we just have to check whether there such a mapping for the address
for which the GOMP_MAP_ATTACH. was requested. And, if there was, the
error diagnostic can be skipped.
Unsurprisingly, this issue occurs in real-world code; it was detected in
a code that distributes work via MPI and for some processes, some bounds
ended up to be zero.
libgomp/ChangeLog:
* target.c (gomp_attach_pointer): Return bool; accept additional
bool to optionally silence the fatal pointee-not-found error.
(gomp_map_vars_internal): If the pointee could not be found,
check whether it was mapped as GOMP_MAP_ZERO_LEN_ARRAY_SECTION.
* libgomp.h (gomp_attach_pointer): Update prototype.
* oacc-mem.c (acc_attach_async, goacc_enter_data_internal): Update
calls.
* testsuite/libgomp.c/target-map-zero-sized.c: New test.
* testsuite/libgomp.c/target-map-zero-sized-2.c: New test.
* testsuite/libgomp.c/target-map-zero-sized-3.c: New test.
While the tests checked whether the CUDA/HIP runtime is available
before processing them, the execution was then done unconditionally,
leading to FAIL when the default device was the host (or the wrong
offload device).
Now the test is only executed ('run') when the default device is an
Nvidia or AMD GPU (depending on the test case, cf. the test file name).
Otherwise, only a 'link' test is done. (Except when the effective-target
check cannot find the runtime lib - then the test is skipped [as before].)
Note: The cublas/hipblas tests use variant functions and iterate over
all devices, such that the cublas or hipblas, respectively, is only
called when the active device is an AMD or Nvidia device, respectively,
while for the host and other device types the fallback is called.
libgomp/ChangeLog:
* testsuite/libgomp.c/interop-cuda-full.c: Use 'link' instead
of 'run' when the default device is "! offload_device_nvptx".
* testsuite/libgomp.c/interop-cuda-libonly.c: Likewise.
* testsuite/libgomp.c/interop-hip-nvidia-full.c: Likewise.
* testsuite/libgomp.c/interop-hip-nvidia-no-headers.c: Likewise.
* testsuite/libgomp.c/interop-hip-nvidia-no-hip-header.c: Likewise.
* testsuite/libgomp.fortran/interop-hip-nvidia-full.F90: Likewise.
* testsuite/libgomp.fortran/interop-hip-nvidia-no-module.F90: Likewise.
* testsuite/libgomp.c/interop-hip-amd-full.c: Use 'link' instead
of 'run' when the default device is "! offload_device_gcn".
* testsuite/libgomp.c/interop-hip-amd-no-hip-header.c: Likewise.
* testsuite/libgomp.fortran/interop-hip-amd-full.F90: Likewise.
* testsuite/libgomp.fortran/interop-hip-amd-no-module.F90: Likewise.
When host memory is device accessible - independent whether mapping is done or
not (i.e. self map), the 'vtab' pointer becomes accessible, which stores the
dynamic type's type and size information.
In principle, we want to test: USM available but mapping is still done, but
as there is no simple + reliable not-crashing way to test for this, those
checks are skipped in the (pre)existing test file map-alloc-comp-9.f90.
Or rather: those are only active with self-maps, which is currently only true
for the host.
This commit adds map-alloc-comp-9-usm.f90 which runs the same test with
'omp requires unified_shared_memory'. While OpenMP permits both actual
mapping and self maps with this flag, it in theory covers the missing cases.
However, currently, GCC always uses self maps with USM. Still, having a
device-run self-maps check is better than nothing, even if it misses the
most interesting case.
libgomp/ChangeLog:
* testsuite/libgomp.fortran/map-alloc-comp-9.f90: Process differently
when USE_USM_REQUIREMENT is set.
* testsuite/libgomp.fortran/map-alloc-comp-9-usm.f90: New test.
Thomas Schwinge [Mon, 5 May 2025 08:19:30 +0000 (10:19 +0200)]
'libgomp.c/interop-hsa.c': GCN offloading only
Fix-up for commit 8d84ea28510054fbbb8a2b7441916bd75e29163f
"OpenMP, GCN: Add interop-hsa testcase", which added 'libgomp.c/interop-hsa.c'.
If nvptx offloading compilation is enabled in addition to GCN, the former ICEs:
Tobias Burnus [Thu, 1 May 2025 15:39:42 +0000 (15:39 +0000)]
OpenMP: Restore lost Fortran testcase for 'omp allocate'
This testcase, which is present on the OG13 and OG14 branches, was
overlooked when the Fortran support for 'omp allocate' was added to
mainline (commit d4b6d147920b93297e621124a99ed01e7e310d92 from
December 2023).
libgomp/ChangeLog
* testsuite/libgomp.fortran/allocate-8a.f90: New test.
libgomp/testsuite: Fix hip_header_nvidia check, add workaround to test
This is all about using the AMD's HIP header files with
__HIP_PLATFORM_NVIDIA__ defined, i.e. HIP with Nvidia/CUDA; in that case,
HIP is a thin layer on top of CUDA.
First, the check_effective_target_gomp_hip_header_nvidia check failed;
to fix it, -Wno-deprecated-declarations was added - and likewise to the
two affected testcases that actually used the HIP headers on Nvidia.
Doing so, the HIP tested was successful but the HIP-BLAS one showed two
issues:
* One seems to be related to include search paths as the HIP header uses
#include "library_types.h" to include that CUDA header. Seemingly, it
tried to included (again) the HIP header hip/library_types.h, not the
CUDA one. I guess, some tweaking of -isystem vs. -I could have
prevented this, but the simpler workaround was to just explicitly
include the CUDA one before the HIP header files.
* Once done, everything compiled but linking failed as the association
between three HIP-BLAS functions and their CUDA-BLAS ones did not
work. Solution: Just add three #define for mapping them.
libgomp/ChangeLog:
* testsuite/lib/libgomp.exp
(check_effective_target_gomp_hip_header_nvidia): Compile with
"-Wno-deprecated-declarations".
* testsuite/libgomp.c/interop-hip-nvidia-full.c: Likewise.
* testsuite/libgomp.c/interop-hipblas-nvidia-full.c: Likewise.
* testsuite/libgomp.c/interop-hipblas.h: Add workarounds
when using the HIP headers with __HIP_PLATFORM_NVIDIA__.
Add checks for nowait/depend and for checks that the returned
CUDA, CUDA_DRIVER and HIP interop objects actually work.
While the CUDA/CUDA_DRIVER ones are only for Nvidia GPUs, HIP
works on both AMD and Nvidia GPUs; on Nvidia GPUs, it is a
very thin wrapper around CUDA.
For Fortran, only a HIP test has been added - using hipfort.
While libgomp.c-c++-common/interop-2.c always works - even without
GPU - and checks for depend / nowait, all others require that
runtime libraries are found at link (and execution) time:
For Nvidia GPUs, libcuda + libcudart or libcublas,
For AMD GPUs, libamdhip64 or libhipblas.
The header files and hipfort modules do not need to be present as a
fallback has been implemented, but if they are, they get used.
Due to the combinations, the basic 1x C/C++, 4x C and 1x Fortran tests
yield 1x C/C++, 14x C and 4 Fortran run-test files.
libgomp/ChangeLog:
* testsuite/lib/libgomp.exp (check_effective_target_openacc_cublas,
check_effective_target_openacc_cudart): Update description as
the check requires more.
(check_effective_target_openacc_libcuda,
check_effective_target_openacc_libcublas,
check_effective_target_openacc_libcudart,
check_effective_target_gomp_hip_header_amd,
check_effective_target_gomp_hip_header_nvidia,
check_effective_target_gomp_hipfort_module,
check_effective_target_gomp_libamdhip64,
check_effective_target_gomp_libhipblas): New.
* testsuite/libgomp.c-c++-common/interop-2.c: New test.
* testsuite/libgomp.c/interop-cublas-full.c: New test.
* testsuite/libgomp.c/interop-cublas-libonly.c: New test.
* testsuite/libgomp.c/interop-cuda-full.c: New test.
* testsuite/libgomp.c/interop-cuda-libonly.c: New test.
* testsuite/libgomp.c/interop-hip-amd-full.c: New test.
* testsuite/libgomp.c/interop-hip-amd-no-hip-header.c: New test.
* testsuite/libgomp.c/interop-hip-nvidia-full.c: New test.
* testsuite/libgomp.c/interop-hip-nvidia-no-headers.c: New test.
* testsuite/libgomp.c/interop-hip-nvidia-no-hip-header.c: New test.
* testsuite/libgomp.c/interop-hip.h: New test.
* testsuite/libgomp.c/interop-hipblas-amd-full.c: New test.
* testsuite/libgomp.c/interop-hipblas-amd-no-hip-header.c: New test.
* testsuite/libgomp.c/interop-hipblas-nvidia-full.c: New test.
* testsuite/libgomp.c/interop-hipblas-nvidia-no-headers.c: New test.
* testsuite/libgomp.c/interop-hipblas-nvidia-no-hip-header.c: New test.
* testsuite/libgomp.c/interop-hipblas.h: New test.
* testsuite/libgomp.fortran/interop-hip-amd-full.F90: New test.
* testsuite/libgomp.fortran/interop-hip-amd-no-module.F90: New test.
* testsuite/libgomp.fortran/interop-hip-nvidia-full.F90: New test.
* testsuite/libgomp.fortran/interop-hip-nvidia-no-module.F90: New test.
* testsuite/libgomp.fortran/interop-hip.h: New test.
Tobias Burnus [Wed, 14 May 2025 18:06:49 +0000 (20:06 +0200)]
OpenMP: Fix mapping of zero-sized arrays with non-literal size: map(var[:n]), n = 0
For map(ptr[:0]), the used map kind is GOMP_MAP_ATTACH_ZERO_LENGTH_ARRAY_SECTION
and it is permitted that 'ptr' does not exist. 'ptr' is set to the device
pointee if it exists or to the host value otherwise.
For map(ptr[:3]), the variable is first mapped and then ptr is updated to point
to the just-mapped device data; the attachment uses GOMP_MAP_ATTACH.
For map(ptr[:n]), generates always a GOMP_MAP_ATTACH, but when n == 0, it
was failing with:
"pointer target not mapped for attach"
The solution is not to fail but first to check whether it was mapped before.
It turned out that for the mapping part, GCC adds a run-time check whether
n == 0 - and uses GOMP_MAP_ZERO_LEN_ARRAY_SECTION for the mapping.
Thus, we just have to check whether there such a mapping for the address
for which the GOMP_MAP_ATTACH. was requested. And, if there was, the
error diagnostic can be skipped.
Unsurprisingly, this issue occurs in real-world code; it was detected in
a code that distributes work via MPI and for some processes, some bounds
ended up to be zero.
libgomp/ChangeLog:
* target.c (gomp_attach_pointer): Return bool; accept additional
bool to optionally silence the fatal pointee-not-found error.
(gomp_map_vars_internal): If the pointee could not be found,
check whether it was mapped as GOMP_MAP_ZERO_LEN_ARRAY_SECTION.
* libgomp.h (gomp_attach_pointer): Update prototype.
* oacc-mem.c (acc_attach_async, goacc_enter_data_internal): Update
calls.
* testsuite/libgomp.c/target-map-zero-sized.c: New test.
* testsuite/libgomp.c/target-map-zero-sized-2.c: New test.
* testsuite/libgomp.c/target-map-zero-sized-3.c: New test.
Alpha: Fix base block alignment calculation regression
In determination of base block alignment we only examine a COMPONENT_REF
tree node at hand without ever checking if its ultimate alignment has
been reduced by the combined offset going back to the outermost object.
Consequently cases have been observed where quadword accesses have been
produced for a memory location referring a nested struct member only
aligned to the longword boundary, causing emulation to trigger.
Address this issue by recursing into COMPONENT_REF tree nodes until the
outermost one has been reached, which is supposed to be a MEM_REF one,
accumulating the offset as we go, fixing a commit e0dae4da4c45 ("Alpha:
Also use tree information to get base block alignment") regression.
Bail out and refrain from using tree information for alignment if we end
up at something different or we are unable to calculate the offset at
any point.
gcc/
* config/alpha/alpha.cc
(alpha_get_mem_rtx_alignment_and_offset): Recurse into
COMPONENT_REF nodes.
gcc/testsuite/
* gcc.target/alpha/memcpy-nested-offset-long.c: New file.
* gcc.target/alpha/memcpy-nested-offset-quad.c: New file.
Patrick Palka [Thu, 15 May 2025 15:07:53 +0000 (11:07 -0400)]
c++: unifying specializations of non-primary tmpls [PR120161]
Here unification of P=Wrap<int>::type, A=Wrap<long>::type wrongly
succeeds ever since r14-4112 which made the RECORD_TYPE case of unify
no longer recurse into template arguments for non-primary templates
(since they're a non-deduced context) and so the int/long mismatch that
makes the two types distinct goes unnoticed.
In the case of (comparing specializations of) a non-primary template,
unify should still go on to compare the types directly before returning
success.
PR c++/120161
gcc/cp/ChangeLog:
* pt.cc (unify) <case RECORD_TYPE>: When comparing specializations
of a non-primary template, still perform a type comparison.
While the tests checked whether the CUDA/HIP runtime is available
before processing them, the execution was then done unconditionally,
leading to FAIL when the default device was the host (or the wrong
offload device).
Now the test is only executed ('run') when the default device is an
Nvidia or AMD GPU (depending on the test case, cf. the test file name).
Otherwise, only a 'link' test is done. (Except when the effective-target
check cannot find the runtime lib - then the test is skipped [as before].)
Note: The cublas/hipblas tests use variant functions and iterate over
all devices, such that the cublas or hipblas, respectively, is only
called when the active device is an AMD or Nvidia device, respectively,
while for the host and other device types the fallback is called.
libgomp/ChangeLog:
* testsuite/libgomp.c/interop-cuda-full.c: Use 'link' instead
of 'run' when the default device is "! offload_device_nvptx".
* testsuite/libgomp.c/interop-cuda-libonly.c: Likewise.
* testsuite/libgomp.c/interop-hip-nvidia-full.c: Likewise.
* testsuite/libgomp.c/interop-hip-nvidia-no-headers.c: Likewise.
* testsuite/libgomp.c/interop-hip-nvidia-no-hip-header.c: Likewise.
* testsuite/libgomp.fortran/interop-hip-nvidia-full.F90: Likewise.
* testsuite/libgomp.fortran/interop-hip-nvidia-no-module.F90: Likewise.
* testsuite/libgomp.c/interop-hip-amd-full.c: Use 'link' instead
of 'run' when the default device is "! offload_device_gcn".
* testsuite/libgomp.c/interop-hip-amd-no-hip-header.c: Likewise.
* testsuite/libgomp.fortran/interop-hip-amd-full.F90: Likewise.
* testsuite/libgomp.fortran/interop-hip-amd-no-module.F90: Likewise.
OpenMP: need_device_ptr and need_device_addr support for adjust_args
This patch adds support for the "need_device_addr" modifier to the
"adjust args" clause for the "declare variant" directive, and
extends/re-works the support for "need_device_ptr" as well.
This patch builds on waffl3x's recently posted patch, "OpenMP: C/C++
adjust-args numeric ranges", here.
In C++, "need_device_addr" supports mapping reference arguments to
device pointers. In Fortran, it similarly supports arguments passed
by reference, the default for the language, in contrast to
"need_device_ptr" which is used to map arguments of c_ptr type. The
C++ support is straightforward, but Fortran has some additional
wrinkles involving arrays passed by descriptor (a new descriptor must
be constructed with a pointer to the array data which is the only part
mapped to the device), plus special cases for passing optional
arguments and a whole array instead of a reference to its first element.
gcc/fortran/ChangeLog
* trans-openmp.cc (gfc_trans_omp_declare_variant): Disallow
polymorphic and optional arguments with need_device_addr for now, but
don't reject need_device_addr entirely.
gcc/ChangeLog
* gimplify.cc (modify_call_for_omp_dispatch): Rework logic for
need_device_ptr and need_device_addr adjustments.
gcc/testsuite/ChangeLog
* c-c++-common/gomp/adjust-args-10.c: Ignore the new sorry since the
lack of proper diagnostic is already xfail'ed.
* g++.dg/gomp/adjust-args-1.C: Adjust output patterns.
* g++.dg/gomp/adjust-args-17.C: New.
* gcc.dg/gomp/adjust-args-3.c: New.
* gfortran.dg/gomp/adjust-args-14.f90: Don't expect this to fail now.
libgomp/ChangeLog
* libgomp.texi: Mark need_device_addr as supported.
* testsuite/libgomp.c++/need-device-ptr.C: New.
* testsuite/libgomp.c-c++-common/dispatch-3.c: New.
* testsuite/libgomp.fortran/adjust-args-array-descriptor.f90: New.
* testsuite/libgomp.fortran/need-device-ptr.f90: New.
waffl3x [Mon, 5 May 2025 20:20:00 +0000 (20:20 +0000)]
OpenMP: C/C++ adjust-args numeric ranges
Add support for OpenMP parameter-lists in an adjust_args clause, more
specifically, numeric ranges and parameter indices. Many bugs are also
fixed along the way. Most of the fixes rely on the changes to handling of
clause arguments and can't reasonably be split out, while most of the
changes that fix PR119602 came as a side effect of properly handling numeric
ranges with dependent bounds so it makes sense to include it here.
Variadic arguments with an incorrect type are not currently diagnosed, but
handling for them is otherwise functional. It is unclear how references to
pointers are supposed to be handled, so for now we sorry for that case.
* c-parser.cc (c_omp_numeric_ranges_always_overlap): New function.
(c_parser_omp_parm_list): New function.
(c_finish_omp_declare_variant): Use c_parser_omp_parm_list instead
of c_parser_omp_variable_list. Refactor, change format of
"omp declare variant variant args" attribute.
gcc/cp/ChangeLog:
PR c++/119659
PR c++/118859
PR c++/119601
PR c++/119602
PR c++/119775
* cp-tree.h (finish_omp_parm_list): New declaration.
(finish_omp_adjust_args): New declaration.
* decl.cc (omp_declare_variant_finalize_one): Refactor and change
attribute unpacking, use finish_omp_parm_list and
finish_omp_adjust_args, refactor append_args diagnostics, add
nbase_parms to append_args attribute, remove special handling for
member functions.
* parser.cc (cp_parser_direct_declarator): Don't pass parms.
(cp_parser_late_return_type_opt): Remove parms parameter.
(cp_parser_omp_parm_list): New function.
(cp_finish_omp_declare_variant): Remove parms parameter.
Add NULL_TREE instead of nbase_args to append_args_tree. Refactor,
use cp_parser_omp_parm_list not cp_parser_omp_var_list_no_open,
handle "need_device_addr" and remove handling and diagnostics of
parm list arguments that are done too early. Change format of
unnamed variant attribute.
(cp_parser_late_parsing_omp_declare_simd): Remove parms parameter.
* pt.cc (tsubst_attribute): Copy "omp declare variant base" nodes,
substitute parm list numeric range bounds.
* semantics.cc (finish_omp_parm_list): New function.
(finish_omp_adjust_args): New function.
gcc/fortran/ChangeLog:
* trans-openmp.cc (gfc_trans_omp_declare_variant): Change format of
"omp declare variant variant args" attribute.
gcc/ChangeLog:
* gimplify.cc (modify_call_for_omp_dispatch): Refactor and change
attribute unpacking. For adjust_args variadic functions, expand
numeric ranges with relative bounds. Refactor argument adjustment.
libgomp/ChangeLog:
* libgomp.texi: Set 'adjust args' variadic arguments support to Y.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/pr118579.c: Change error text.
* g++.dg/gomp/adjust-args-1.C: Fix error text, add dg-* directives.
* g++.dg/gomp/adjust-args-2.C: Add dg-* directives.
* g++.dg/gomp/append-args-1.C: Add dg-* directives.
* gcc.dg/gomp/adjust-args-1.c: Fix error text, add dg-* directives.
* gcc.dg/gomp/append-args-1.c: Fix error text, add dg-* directives.
* c-c++-common/gomp/adjust-args-7.c: New test.
* c-c++-common/gomp/adjust-args-8.c: New test.
* c-c++-common/gomp/adjust-args-9.c: New test.
* c-c++-common/gomp/adjust-args-10.c: New test.
* c-c++-common/gomp/adjust-args-11.c: New test.
* c-c++-common/gomp/adjust-args-12.c: New test.
* c-c++-common/gomp/adjust-args-13.c: New test.
* c-c++-common/gomp/adjust-args-14.c: New test.
* c-c++-common/gomp/adjust-args-15.c: New test.
* g++.dg/gomp/adjust-args-5.C: New test.
* g++.dg/gomp/adjust-args-6.C: New test.
* g++.dg/gomp/adjust-args-7.C: New test.
* g++.dg/gomp/adjust-args-8.C: New test.
* g++.dg/gomp/adjust-args-9.C: New test.
* g++.dg/gomp/adjust-args-10.C: New test.
* g++.dg/gomp/adjust-args-11.C: New test.
* g++.dg/gomp/adjust-args-12.C: New test.
* g++.dg/gomp/adjust-args-13.C: New test.
* g++.dg/gomp/adjust-args-14.C: New test.
* g++.dg/gomp/adjust-args-15.C: New test.
* g++.dg/gomp/adjust-args-16.C: New test.
* g++.dg/gomp/append-args-9.C: New test.
* g++.dg/gomp/append-args-10.C: New test.
* g++.dg/gomp/append-args-11.C: New test.
* g++.dg/gomp/append-args-omp-interop-t.h: New header.
Andrew Pinski [Mon, 21 Apr 2025 22:32:26 +0000 (22:32 +0000)]
GCN: Properly switch sections in 'gcn_hsa_declare_function_name' [PR119737]
There are GCN/C++ target as well as offloading codes, where the hard-coded
section names in 'gcn_hsa_declare_function_name' do not fit, and assembly thus
fails:
LLVM ERROR: Size expression must be absolute.
This commit progresses GCN target:
[-FAIL: g++.dg/init/call1.C -std=gnu++17 (internal compiler error: Aborted signal terminated program as)-]
[-FAIL:-]{+PASS:+} g++.dg/init/call1.C -std=gnu++17 (test for excess errors)
[-UNRESOLVED:-]{+PASS:+} g++.dg/init/call1.C -std=gnu++17 [-compilation failed to produce executable-]{+execution test+}
[-FAIL: g++.dg/init/call1.C -std=gnu++26 (internal compiler error: Aborted signal terminated program as)-]
[-FAIL:-]{+PASS:+} g++.dg/init/call1.C -std=gnu++26 (test for excess errors)
[-UNRESOLVED:-]{+PASS:+} g++.dg/init/call1.C -std=gnu++26 [-compilation failed to produce executable-]{+execution test+}
UNSUPPORTED: g++.dg/init/call1.C -std=gnu++98: exception handling not supported
..., and GCN offloading:
[-XFAIL: libgomp.c++/target-exceptions-throw-1.C (internal compiler error: Aborted signal terminated program as)-]
[-XFAIL: libgomp.c++/target-exceptions-throw-1.C PR119737 at line 7 (test for bogus messages, line )-]
[-XFAIL:-]{+PASS:+} libgomp.c++/target-exceptions-throw-1.C (test for excess errors)
[-UNRESOLVED:-]{+PASS:+} libgomp.c++/target-exceptions-throw-1.C [-compilation failed to produce executable-]{+execution test+}
{+PASS: libgomp.c++/target-exceptions-throw-1.C output pattern test+}
[-XFAIL: libgomp.c++/target-exceptions-throw-2.C (internal compiler error: Aborted signal terminated program as)-]
[-XFAIL: libgomp.c++/target-exceptions-throw-2.C PR119737 at line 7 (test for bogus messages, line )-]
[-XFAIL:-]{+PASS:+} libgomp.c++/target-exceptions-throw-2.C (test for excess errors)
[-UNRESOLVED:-]{+PASS:+} libgomp.c++/target-exceptions-throw-2.C [-compilation failed to produce executable-]{+execution test+}
{+PASS: libgomp.c++/target-exceptions-throw-2.C output pattern test+}
[-XFAIL: libgomp.oacc-c++/exceptions-throw-1.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 (internal compiler error: Aborted signal terminated program as)-]
[-XFAIL: libgomp.oacc-c++/exceptions-throw-1.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 PR119737 at line 7 (test for bogus messages, line )-]
[-XFAIL:-]{+PASS:+} libgomp.oacc-c++/exceptions-throw-1.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 (test for excess errors)
[-UNRESOLVED:-]{+PASS:+} libgomp.oacc-c++/exceptions-throw-1.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 [-compilation failed to produce executable-]{+execution test+}
{+PASS: libgomp.oacc-c++/exceptions-throw-1.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 output pattern test+}
[-XFAIL: libgomp.oacc-c++/exceptions-throw-2.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 (internal compiler error: Aborted signal terminated program as)-]
[-XFAIL: libgomp.oacc-c++/exceptions-throw-2.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 PR119737 at line 9 (test for bogus messages, line )-]
[-XFAIL:-]{+PASS:+} libgomp.oacc-c++/exceptions-throw-2.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 (test for excess errors)
[-UNRESOLVED:-]{+PASS:+} libgomp.oacc-c++/exceptions-throw-2.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 [-compilation failed to produce executable-]{+execution test+}
{+PASS: libgomp.oacc-c++/exceptions-throw-2.C -DACC_DEVICE_TYPE_radeon=1 -DACC_MEM_SHARED=0 -foffload=amdgcn-amdhsa -O2 output pattern test+}
Thomas Schwinge [Tue, 22 Apr 2025 11:41:22 +0000 (13:41 +0200)]
Adjust 'libgomp.c++/target-exceptions-pr118794-1.C' for 'targetm.arm_eabi_unwinder' [PR118794]
Fix-up for commit aa3e72f943032e5f074b2bd2fd06d130dda8760b
"Add test cases for exception handling constructs in dead code for GCN, nvptx target and OpenMP 'target' offloading [PR118794]":
we need to adjust for configurations with 'targetm.arm_eabi_unwinder', as per:
..., which for ARM is conditional to '#if ARM_UNWIND_INFO' (defined in
'gcc/config/arm/bpabi.h', used for various GCC configurations), and for
C6x unconditional.