aboutsummaryrefslogtreecommitdiff
AgeCommit message (Collapse)AuthorFilesLines
2025-05-27libstdc++: Fix vector(from_range_t, R&&) for exceptions [PR120367]Jonathan Wakely2-1/+24
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. Reviewed-by: Patrick Palka <ppalka@redhat.com> (cherry picked from commit 04f2be72b1deecd6c6d454e000cfc0cb16db957c)
2025-05-27AVR: target/120442 - Support f7_fdim / fdiml in LibF7.Georg-Johann Lay6-10/+39
Add Support for fdiml. PR target/120442 libgcc/config/avr/libf7/ * libf7-common.mk (LIBF_C_PARTS, m_ddd): Add fdim. * libf7.h (f7_fdim): New proto. * libf7.c (f7_fdim): New function. * f7renames.sh (f7_fdim): Add rename. * f7-wraps.h: Rebuild * f7-renames.h: Rebuild (cherry picked from commit 6045de6596d09f3cf7ae3f552b56d7e5df17a899)
2025-05-27AVR: target/120441 - Fix f7_exp for |x| ≥ 512.Georg-Johann Lay1-2/+2
f7_exp limited exponents to 512, but 1023 * ln2 ≈ 709, hence 1024 is a correct limit. libgcc/config/avr/libf7/ PR target/120441 * libf7.c (f7_exp): Limit aa->expo to 10 (not to 9). (cherry picked from commit 672569cee76a1927d14b5eb754a5ff0b9cee1bc8)
2025-05-27Daily bump.GCC Administrator4-1/+40
2025-05-26Merge branch 'releases/gcc-15' into devel/omp/gcc-15Tobias Burnus44-39/+577
Merge up to r15-9731-gd390c7e5bd0349 (26th May 2025)
2025-05-26c-c++-common/gomp/{attrs-,}metadirective-3.c: Fix expected result [PR118694]Tobias Burnus2-6/+8
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. (cherry picked from commit 5d6ed6d604ff949b650e48fa4eaed3ec8b6489c1)
2025-05-26libgomp.c-c++-common/metadirective-1.c: Expect 'error:' for nvptx compile ↵Tobias Burnus1-1/+7
[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. (cherry picked from commit b3d07ec7ac2ccd935a79b29e1a0e2eb16225286a)
2025-05-26OpenMP/C++: Avoid ICE for BIND_EXPR with empty BIND_EXPR_BLOCK [PR120413]Tobias Burnus2-4/+26
PR c++/120413 gcc/cp/ChangeLog: * semantics.cc (finish_omp_target_clauses_r): Handle BIND_EXPR with empty BIND_EXPR_BLOCK. gcc/testsuite/ChangeLog: * g++.dg/gomp/target-4.C: New test. (cherry picked from commit 45b849d05b733a25ec7ce612229084b8f4b86d3d)
2025-05-26Daily bump.GCC Administrator2-1/+8
2025-05-25MicroBlaze does not support speculative execution (CVE-2017-5753)Michael J. Eager1-0/+4
gcc/ PR target/86772 Tracking CVE-2017-5753 * config/microblaze/microblaze.cc (TARGET_HAVE_SPECULATION_SAFE_VALUE): Define to speculation_save_value_not_needed
2025-05-25Daily bump.GCC Administrator1-1/+1
2025-05-24Daily bump.GCC Administrator3-1/+47
2025-05-24c++/modules: Fix merge of TLS init functions [PR120363]Nathaniel Shead10-3/+82
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)
2025-05-24c++/modules: Fix stream-in of member using-decls [PR120414]Nathaniel Shead3-1/+25
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)
2025-05-23Daily bump.GCC Administrator3-1/+23
2025-05-22Fortran: default-initialization and functions returning derived type [PR85750]Harald Anlauf7-13/+134
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. gcc/testsuite/ChangeLog: * gfortran.dg/alloc_comp_auto_array_3.f90: Adjust scan counts. * gfortran.dg/alloc_comp_class_3.f03: Remove bogus warnings. * gfortran.dg/alloc_comp_class_4.f03: Likewise. * gfortran.dg/allocate_with_source_14.f03: Adjust scan count. * gfortran.dg/derived_constructor_comps_6.f90: Likewise. * gfortran.dg/derived_result_5.f90: New test. (cherry picked from commit d31ab498b12ebbe4f50acb2aa240ff92c73f310c)
2025-05-22ChangeLog.omp bumpThomas Schwinge3-0/+52
2025-05-22'TYPE_EMPTY_P' vs. code offloading [PR120308]Thomas Schwinge6-3/+61
We've got 'gcc/stor-layout.cc:finalize_type_size': /* 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'.) PR lto/120308 gcc/ * lto-streamer-out.cc (hash_tree): Don't handle 'TYPE_EMPTY_P' for 'lto_stream_offload_p'. * tree-streamer-in.cc (unpack_ts_type_common_value_fields): Likewise. * tree-streamer-out.cc (pack_ts_type_common_value_fields): Likewise. libgomp/ * testsuite/libgomp.oacc-c-c++-common/abi-struct-1.c: Add empty structure testing. gcc/testsuite/ * gcc.target/nvptx/abi-struct-arg.c: Add empty structure testing. * gcc.target/nvptx/abi-struct-ret.c: Likewise. (cherry picked from commit 9063810c86beee6274d745b91d8fb43a81c9683e)
2025-05-22Add 'libgomp.c-c++-common/target-abi-struct-1-O0.c', ↵Thomas Schwinge2-0/+99
'libgomp.oacc-c-c++-common/abi-struct-1.c' libgomp/ * testsuite/libgomp.c-c++-common/target-abi-struct-1-O0.c: New. * testsuite/libgomp.oacc-c-c++-common/abi-struct-1.c: Likewise. (cherry picked from commit 45efda05c47f770a617b44cf85713a696bcf0384)
2025-05-22GCN, nvptx offloading: Restrain 'WARNING: program timed out.' while in ↵Thomas Schwinge6-0/+18
'dynamic_cast'" [PR119692] PR target/119692 libgomp/ * testsuite/libgomp.c++/pr119692-1-4.C: '{ dg-timeout 10 }'. * testsuite/libgomp.c++/pr119692-1-5.C: Likewise. * testsuite/libgomp.c++/target-exceptions-bad_cast-1.C: Likewise. * testsuite/libgomp.c++/target-exceptions-bad_cast-2.C: Likewise. * testsuite/libgomp.oacc-c++/exceptions-bad_cast-1.C: Likewise. * testsuite/libgomp.oacc-c++/exceptions-bad_cast-2.C: Likewise. (cherry picked from commit b5f48e7872db30b8f174cb2c497868a358bf75d6)
2025-05-22ChangeLog.omp bumpTobias Burnus2-1/+8
2025-05-22git_repository.py: Fix handling of --last-commit with mergesTobias Burnus2-16/+21
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.
2025-05-22Daily bump.GCC Administrator1-1/+1
2025-05-21Daily bump.GCC Administrator6-1/+103
2025-05-20Fortran: fix passing of inquiry ref of complex array to TRANSFER [PR102891]Harald Anlauf2-1/+53
PR fortran/102891 gcc/fortran/ChangeLog: * dependency.cc (gfc_ref_needs_temporary_p): Within an array reference, inquiry references of complex variables generally need a temporary. gcc/testsuite/ChangeLog: * gfortran.dg/transfer_array_subref.f90: New test. (cherry picked from commit 94fa992b60e53dcf807fc7055ab606d828b931d8)
2025-05-20tree-sra: Do not create stores into const aggregates (PR111873)Martin Jambor4-1/+52
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. (cherry picked from commit 9d039eff453f777c58642ff16178c1ce2a4be6ab)
2025-05-20ipa: Dump cgraph_node UID instead of order into ipa-clones dump fileMartin Jambor6-8/+8
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. Co-Authored-By: Michal Jires <mjires@suse.cz> (cherry picked from commit 9fa534f0831892393885e64596a0d6ca8c4078b6)
2025-05-20Merge branch 'releases/gcc-15' into devel/omp/gcc-15Tobias Burnus136-600/+3958
Merge up to r15-9715-g911cfea5e59798 (May 20, 2025).
2025-05-20libstdc++: Fix incorrect links to archived SGI STL docsJonathan Wakely11-24/+24
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. libstdc++-v3/ChangeLog: * doc/xml/faq.xml: Update URL for archived SGI STL docs. * doc/xml/manual/containers.xml: Likewise. * doc/xml/manual/extensions.xml: Likewise. * doc/xml/manual/using.xml: Likewise. * doc/xml/manual/utilities.xml: Likewise. * doc/html/*: Regenerate. (cherry picked from commit 501e6e786652748ff0ad9a322f74b9b47970031f)
2025-05-20c++/modules: Fix ICE on merge of instantiation with partial spec [PR120013]Nathaniel Shead6-11/+72
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)
2025-05-20c++/modules: Always mark tinfo vars as TREE_ADDRESSABLE [PR120350]Nathaniel Shead3-0/+16
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)
2025-05-20Daily bump.GCC Administrator4-1/+186
2025-05-19libstdc++: Fix some Clang -Wsystem-headers warnings in <ranges>Jonathan Wakely1-6/+6
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. Reviewed-by: Tomasz Kamiński <tkaminsk@redhat.com> (cherry picked from commit 1197f896ae5558f27baa929a10f66447aaafb681)
2025-05-19libstdc++: Fix std::format of chrono::local_days with {} [PR120293]Jonathan Wakely2-0/+6
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. Reviewed-by: Tomasz Kamiński <tkaminsk@redhat.com> (cherry picked from commit 1ed7585bf60ba9940ca5dc6d2c72dba86eea7b4d)
2025-05-19libstdc++: Fix dangling pointer in fs::path::operator+=(*this) [PR120029]Jonathan Wakely3-0/+156
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. (cherry picked from commit a067cbcdcc5f599a2b7d607e89674533d23c652d)
2025-05-19libstdc++: Fix std::format_kind primary template for Clang [PR120190]Jonathan Wakely2-10/+24
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)
2025-05-19OpenMP/Fortran: Fix allocatable-component mapping of derived-type array compsTobias Burnus2-0/+95
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. (cherry picked from commit f99017c3125f4400cf6a098cf5b33d32fe3e6645)
2025-05-19OpenMP: Fix mapping of zero-sized arrays with non-literal size: ↵Tobias Burnus6-16/+288
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. (cherry picked from commit 814e29e390b1e9253f9a38e0d84f5ebe5de0c13e)
2025-05-19libgomp.{c,fortran}/interop-{hip,cuda}: Fix dg-run target selectionTobias Burnus11-0/+33
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. (cherry picked from commit 94e63410474a36655e1800387eabd73a6f930048)
2025-05-19libgomp.fortran/map-alloc-comp-9{,-usm}.f90: Add unified_shared_memory variantTobias Burnus2-0/+30
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. (cherry picked from commit 9565076f9b810541aeb63cb621d694326aa12216)
2025-05-19'libgomp.c/interop-hsa.c': GCN offloading onlyThomas Schwinge1-1/+3
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: during RTL pass: final [...]/libgomp.c/interop-hsa.c: In function 'get_kernel_ptr': [...]/libgomp.c/interop-hsa.c:34:1: internal compiler error: RTL check: expected code 'subreg', have 'reg' in nvptx_print_operand, at config/nvptx/nvptx.cc:3082 0x1ccdb96 internal_error(char const*, ...) [...]/gcc/diagnostic-global-context.cc:517 0x7446c3 rtl_check_failed_code1(rtx_def const*, rtx_code, char const*, int, char const*) [...]/gcc/rtl.cc:770 0x7fa533 nvptx_print_operand [...]/gcc/config/nvptx/nvptx.cc:3082 0xb25f34 output_operand(rtx_def*, int) [...]/gcc/final.cc:3641 0xb26f07 output_asm_insn(char const*, rtx_def**) [...]/gcc/final.cc:3534 0xb29d91 output_asm_insn(char const*, rtx_def**) [...]/gcc/final.cc:2639 0xb29d91 final_scan_insn_1 [...]/gcc/final.cc:2642 0xb2a59f final_scan_insn(rtx_insn*, _IO_FILE*, int, int, int*) [...]/gcc/final.cc:2892 0xb2a68c final_1 [...]/gcc/final.cc:1983 0xb2b378 rest_of_handle_final [...]/gcc/final.cc:4250 0xb2b378 execute [...]/gcc/final.cc:4328 Regardless of the issue that nvptx offloading compilation probably shouldn't ICE, the 'asm' insert clearly is valid for GCN only. libgomp/ * testsuite/libgomp.c/interop-hsa.c: GCN offloading only. (cherry picked from commit 85ad0d84fcec720c1d94b9bda9a617ced70ba5d2)
2025-05-19OpenMP: Restore lost Fortran testcase for 'omp allocate'Tobias Burnus1-0/+45
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. (cherry picked from commit 08ce1b9f6707e00089c4d77d2bb82963d531bb1d)
2025-05-19OpenMP, GCN: Add interop-hsa testcaseAndrew Stubbs1-0/+203
This testcase ensures that the interop HSA support is sufficient to run a kernel manually on the same device. libgomp/ChangeLog: * testsuite/libgomp.c/interop-hsa.c: New test. (cherry picked from commit 8d84ea28510054fbbb8a2b7441916bd75e29163f)
2025-05-19libgomp/testsuite: Fix hip_header_nvidia check, add workaround to testTobias Burnus4-4/+16
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__. (cherry picked from commit 8ef0518bce489c4c0c252a0e0c44193c5f7cf777)
2025-05-19libgomp: Add additional OpenMP interop runtime testsTobias Burnus23-2/+1397
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. (cherry picked from commit 515d9be7944e89f5ec4363f9816ad4031ab6394b)
2025-05-19OpenMP: Add libgomp.fortran/target-enter-data-8.f90Tobias Burnus1-0/+532
Add another testcase for Fortran deep mapping of allocatable components. libgomp/ChangeLog: * testsuite/libgomp.fortran/target-enter-data-8.f90: New test. (cherry picked from commit c9a8f2f9d39a317ed67fb47157a995ea03c182d4)
2025-05-19Daily bump.GCC Administrator1-1/+1
2025-05-18Daily bump.GCC Administrator3-1/+20
2025-05-17Fortran: Fix ICE with use of c_associated.Jerry DeLisle3-16/+49
PR fortran/120049 gcc/fortran/ChangeLog: * check.cc (gfc_check_c_associated): Modify checks to avoid ICE and allow use, intrinsic :: iso_c_binding from a separate module file. gcc/testsuite/ChangeLog: * gfortran.dg/pr120049_a.f90: New test. * gfortran.dg/pr120049_b.f90: New test. (cherry picked from commit d0571638a6bad932b226ada98b167fa47a47d838)
2025-05-17Daily bump.GCC Administrator4-1/+22
generated by cgit v1.1 at 2025-09-18 17:00:14 +0000