aboutsummaryrefslogtreecommitdiff
path: root/gcc
AgeCommit message (Collapse)AuthorFilesLines
2023-06-26d: Merge upstream dmd, druntime a45f4e9f43, phobos 106038f2e.Iain Buclaw42-205/+428
D front-end changes: - Import dmd v2.103.1. - Deprecated invalid special token sequences inside token strings. D runtime changes: - Import druntime v2.103.1. Phobos changes: - Import phobos v2.103.1. gcc/d/ChangeLog: * dmd/MERGE: Merge upstream dmd a45f4e9f43. * dmd/VERSION: Bump version to v2.103.1. libphobos/ChangeLog: * libdruntime/MERGE: Merge upstream druntime a45f4e9f43. * src/MERGE: Merge upstream phobos 106038f2e. (cherry picked from commit 55620c7bd31705fe024eb6b6ab517981b57a9296)
2023-06-25Daily bump.GCC Administrator1-1/+1
2023-06-24Daily bump.GCC Administrator4-1/+97
2023-06-23compiler, libgo: support bootstrapping gc compilerIan Lance Taylor2-2/+32
In the Go 1.21 release the package internal/profile imports internal/lazyregexp. That works when bootstrapping with Go 1.17, because that compiler has internal/lazyregep and permits importing it. We also have internal/lazyregexp in libgo, but since it is not installed it is not available for importing. This CL adds internal/lazyregexp to the list of internal packages that are installed for bootstrapping. The Go 1.21, and earlier, releases have a couple of functions in the internal/abi package that are always fully intrinsified. The gofrontend recognizes and intrinsifies those functions as well. However, the gofrontend was also building function descriptors for references to the functions without calling them, which failed because there was nothing to refer to. That is OK for the gc compiler, which guarantees that the functions are only called, not referenced. This CL arranges to not generate function descriptors for these functions. For golang/go#60913 Reviewed-on: https://go-review.googlesource.com/c/gofrontend/+/504798
2023-06-23c++: fix PR110102 backportJason Merrill1-3/+2
The argument for is_xible changed from list to vec in GCC 14. gcc/cp/ChangeLog: * call.cc (maybe_init_list_as_array): Build a list.
2023-06-23c++: init-list of uncopyable type [PR110102]Jason Merrill2-0/+23
The maybe_init_list_as_range optimization is a form of copy elision, but we can only elide well-formed copies. PR c++/110102 gcc/cp/ChangeLog: * call.cc (maybe_init_list_as_array): Check that the element type is copyable. gcc/testsuite/ChangeLog: * g++.dg/cpp0x/initlist-opt1.C: New test.
2023-06-23c++: fix explicit/copy problem [PR109247]Jason Merrill2-0/+46
In the testcase, the user wants the assignment to use the operator= declared in the class, but because [over.match.list] says that explicit constructors are also considered for list-initialization, as affirmed in CWG1228, we end up choosing the implicitly-declared copy assignment operator, using the explicit constructor template for the argument, which is ill-formed. Other implementations haven't implemented CWG1228, so we keep getting bug reports. Discussion in CWG led to the idea for this targeted relaxation: if we use an explicit constructor for the conversion to the argument of a copy or move special member function, that makes the candidate worse than another. DR 2735 PR c++/109247 gcc/cp/ChangeLog: * call.cc (sfk_copy_or_move): New. (joust): Add tiebreaker for explicit conv and copy ctor. gcc/testsuite/ChangeLog: * g++.dg/cpp0x/initlist-explicit3.C: New test.
2023-06-23tree-optimization/110298 - CFG cleanup and stale nb_iterationsRichard Biener2-3/+24
When unrolling we eventually kill nb_iterations info since it may refer to removed SSA names. But we do this only after cleaning up the CFG which in turn can end up accessing it. Fixed by swapping the two. PR tree-optimization/110298 * tree-ssa-loop-ivcanon.cc (tree_unroll_loops_completely): Clear number of iterations info before cleaning up the CFG. * gcc.dg/torture/pr110298.c: New testcase. (cherry picked from commit 916add3bf6e46467e4391e358b11ecfbc4daa275)
2023-06-23debug/110295 - mixed up early/late debug for member DIEsRichard Biener2-1/+21
When we process a scope typedef during early debug creation and we have already created a DIE for the type when the decl is TYPE_DECL_IS_STUB and this DIE is still in limbo we end up just re-parenting that type DIE instead of properly creating a DIE for the decl, eventually picking up the now completed type and creating DIEs for the members. Instead this is currently defered to the second time we come here, when we annotate the DIEs with locations late where now the type DIE is no longer in limbo and we fall through doing the job for the decl. The following makes sure we perform the necessary early tasks for this by continuing with the decl DIE creation after setting a parent for the limbo type DIE. PR debug/110295 * dwarf2out.cc (process_scope_var): Continue processing the decl after setting a parent in case the existing DIE was in limbo. * g++.dg/debug/pr110295.C: New testcase. (cherry picked from commit 963f87f8a65ec82f503ac4334a3da83b0a8a43b2)
2023-06-23middle-end/110055 - avoid CLOBBERing static variablesRichard Biener2-1/+19
The gimplifier can elide initialized constant automatic variables to static storage in which case TARGET_EXPR gimplification needs to avoid emitting a CLOBBER for them since their lifetime is no longer limited. Failing to do so causes spurious dangling-pointer diagnostics on the added testcase for some targets. PR middle-end/110055 * gimplify.cc (gimplify_target_expr): Do not emit CLOBBERs for variables which have static storage duration after gimplifying their initializers. * g++.dg/warn/Wdangling-pointer-pr110055.C: New testcase. (cherry picked from commit 84eec2916fa68cd2e2b3a2cf764f2ba595cce843)
2023-06-23ipa/109983 - (IPA) PTA speedupRichard Biener1-71/+46
This improves the edge avoidance heuristic by re-ordering the topological sort of the graph to make sure the component with the ESCAPED node is processed first. This improves the number of created edges which directly correlates with the number of bitmap_ior_into calls from 141447426 to 239596 and the compile-time from 1083s to 3s. It also improves the compile-time for the related PR109143 from 81s to 27s. I've modernized the topological sorting API on the way as well. PR ipa/109983 PR tree-optimization/109143 * tree-ssa-structalias.cc (struct topo_info): Remove. (init_topo_info): Likewise. (free_topo_info): Likewise. (compute_topo_order): Simplify API, put the component with ESCAPED last so it's processed first. (topo_visit): Adjust. (solve_graph): Likewise. (cherry picked from commit 95e5c38a98cc64a797b1d766a20f8c0d0c807a74)
2023-06-23Daily bump.GCC Administrator1-1/+1
2023-06-22Merge branch 'releases/gcc-13' into devel/omp/gcc-13Tobias Burnus25-98/+445
Merge up to r13-7464-gd116ce761052947efa392ae95c7ca6230147ac29 (22nd June 2023)
2023-06-22Daily bump.GCC Administrator1-1/+1
2023-06-21Daily bump.GCC Administrator3-1/+111
2023-06-20aarch64: Allow compiler to define ls64 builtins [PR110132]Alex Coplan8-39/+100
This patch refactors the ls64 builtins to allow the compiler to define them directly instead of having wrapper functions in arm_acle.h. This should be not only easier to maintain, but it makes two important correctness fixes: - It fixes PR110132, where the builtins ended up getting declared with invisible bindings in the C FE, so the FE ended up synthesizing incompatible implicit definitions for these builtins. - It allows the builtins to be used with LTO, which didn't work previously. We also take the opportunity to add test coverage from C++ for these builtins. gcc/ChangeLog: PR target/110132 * config/aarch64/aarch64-builtins.cc (aarch64_general_simulate_builtin): New. Use it ... (aarch64_init_ls64_builtins): ... here. Switch to declaring public ACLE names for builtins. (aarch64_general_init_builtins): Ensure we invoke the arm_acle.h setup if in_lto_p, just like we do for SVE. * config/aarch64/arm_acle.h: (__arm_ld64b): Delete. (__arm_st64b): Delete. (__arm_st64bv): Delete. (__arm_st64bv0): Delete. gcc/testsuite/ChangeLog: PR target/110132 * lib/target-supports.exp (check_effective_target_aarch64_asm_FUNC_ok): Extend to ls64. * g++.target/aarch64/acle/acle.exp: New. * g++.target/aarch64/acle/ls64.C: New test. * g++.target/aarch64/acle/ls64_lto.C: New test. * gcc.target/aarch64/acle/ls64_lto.c: New test. * gcc.target/aarch64/acle/pr110132.c: New test. (cherry picked from commit 9963029a24f2d2510b82e7106fae3f364da33c5d)
2023-06-20aarch64: Fix wrong code with st64b builtin [PR110100]Alex Coplan3-2/+9
The st64b pattern incorrectly had an output constraint on the register operand containing the destination address for the store, leading to wrong code. This patch fixes that. gcc/ChangeLog: PR target/110100 * config/aarch64/aarch64-builtins.cc (aarch64_expand_builtin_ls64): Use input operand for the destination address. * config/aarch64/aarch64.md (st64b): Fix constraint on address operand. gcc/testsuite/ChangeLog: PR target/110100 * gcc.target/aarch64/acle/pr110100.c: New test. (cherry picked from commit 737a0b749a7bc3e7cb904ea2d4b18dc130514b85)
2023-06-20aarch64: Fix whitespace in ls64 builtin implementation [PR110100]Alex Coplan2-43/+43
The ls64 builtin code was using incorrect GNU style with eight spaces where there should be a tab. Fixed thusly. gcc/ChangeLog: PR target/110100 * config/aarch64/aarch64-builtins.cc (aarch64_init_ls64_builtins_types): Replace eight consecutive spaces with tabs. (aarch64_init_ls64_builtins): Likewise. (aarch64_expand_builtin_ls64): Likewise. * config/aarch64/aarch64.md (ld64b): Likewise. (st64b): Likewise. (st64bv): Likewise (st64bv0): Likewise. (cherry picked from commit 713613541254039a34e1dd8fd4a613a299af1fd6)
2023-06-20amdgcn: implement vector div and mod libfuncsAndrew Stubbs103-7/+1315
Also divmod, but only for scalar modes, for now (because there are no complex int vectors yet). gcc/ChangeLog: * config/gcn/gcn.cc (gcn_expand_divmod_libfunc): New function. (gcn_init_libfuncs): Add div and mod functions for all modes. Add placeholders for divmod functions. (TARGET_EXPAND_DIVMOD_LIBFUNC): Define. libgcc/ChangeLog: * config/gcn/lib2-divmod-di.c: Reimplement like lib2-divmod.c. * config/gcn/lib2-divmod.c: Likewise. * config/gcn/lib2-gcn.h: Add new types and prototypes for all the new vector libfuncs. * config/gcn/t-amdgcn: Add new files. * config/gcn/amdgcn_veclib.h: New file. * config/gcn/lib2-vec_divmod-di.c: New file. * config/gcn/lib2-vec_divmod-hi.c: New file. * config/gcn/lib2-vec_divmod-qi.c: New file. * config/gcn/lib2-vec_divmod.c: New file. gcc/testsuite/ChangeLog: * gcc.dg/tree-ssa/predcom-2.c: Avoid vectors on amdgcn. * gcc.dg/unroll-8.c: Likewise. * gcc.dg/vect/slp-26.c: Change expected results on amdgdn. * lib/target-supports.exp (check_effective_target_vect_int_mod): Add amdgcn. (check_effective_target_divmod): Likewise. * gcc.target/gcn/simd-math-3-16.c: New test. * gcc.target/gcn/simd-math-3-2.c: New test. * gcc.target/gcn/simd-math-3-32.c: New test. * gcc.target/gcn/simd-math-3-4.c: New test. * gcc.target/gcn/simd-math-3-8.c: New test. * gcc.target/gcn/simd-math-3-char-16.c: New test. * gcc.target/gcn/simd-math-3-char-2.c: New test. * gcc.target/gcn/simd-math-3-char-32.c: New test. * gcc.target/gcn/simd-math-3-char-4.c: New test. * gcc.target/gcn/simd-math-3-char-8.c: New test. * gcc.target/gcn/simd-math-3-char-run-16.c: New test. * gcc.target/gcn/simd-math-3-char-run-2.c: New test. * gcc.target/gcn/simd-math-3-char-run-32.c: New test. * gcc.target/gcn/simd-math-3-char-run-4.c: New test. * gcc.target/gcn/simd-math-3-char-run-8.c: New test. * gcc.target/gcn/simd-math-3-char-run.c: New test. * gcc.target/gcn/simd-math-3-char.c: New test. * gcc.target/gcn/simd-math-3-long-16.c: New test. * gcc.target/gcn/simd-math-3-long-2.c: New test. * gcc.target/gcn/simd-math-3-long-32.c: New test. * gcc.target/gcn/simd-math-3-long-4.c: New test. * gcc.target/gcn/simd-math-3-long-8.c: New test. * gcc.target/gcn/simd-math-3-long-run-16.c: New test. * gcc.target/gcn/simd-math-3-long-run-2.c: New test. * gcc.target/gcn/simd-math-3-long-run-32.c: New test. * gcc.target/gcn/simd-math-3-long-run-4.c: New test. * gcc.target/gcn/simd-math-3-long-run-8.c: New test. * gcc.target/gcn/simd-math-3-long-run.c: New test. * gcc.target/gcn/simd-math-3-long.c: New test. * gcc.target/gcn/simd-math-3-run-16.c: New test. * gcc.target/gcn/simd-math-3-run-2.c: New test. * gcc.target/gcn/simd-math-3-run-32.c: New test. * gcc.target/gcn/simd-math-3-run-4.c: New test. * gcc.target/gcn/simd-math-3-run-8.c: New test. * gcc.target/gcn/simd-math-3-run.c: New test. * gcc.target/gcn/simd-math-3-short-16.c: New test. * gcc.target/gcn/simd-math-3-short-2.c: New test. * gcc.target/gcn/simd-math-3-short-32.c: New test. * gcc.target/gcn/simd-math-3-short-4.c: New test. * gcc.target/gcn/simd-math-3-short-8.c: New test. * gcc.target/gcn/simd-math-3-short-run-16.c: New test. * gcc.target/gcn/simd-math-3-short-run-2.c: New test. * gcc.target/gcn/simd-math-3-short-run-32.c: New test. * gcc.target/gcn/simd-math-3-short-run-4.c: New test. * gcc.target/gcn/simd-math-3-short-run-8.c: New test. * gcc.target/gcn/simd-math-3-short-run.c: New test. * gcc.target/gcn/simd-math-3-short.c: New test. * gcc.target/gcn/simd-math-3.c: New test. * gcc.target/gcn/simd-math-4-char-run.c: New test. * gcc.target/gcn/simd-math-4-char.c: New test. * gcc.target/gcn/simd-math-4-long-run.c: New test. * gcc.target/gcn/simd-math-4-long.c: New test. * gcc.target/gcn/simd-math-4-run.c: New test. * gcc.target/gcn/simd-math-4-short-run.c: New test. * gcc.target/gcn/simd-math-4-short.c: New test. * gcc.target/gcn/simd-math-4.c: New test. * gcc.target/gcn/simd-math-5-16.c: New test. * gcc.target/gcn/simd-math-5-32.c: New test. * gcc.target/gcn/simd-math-5-4.c: New test. * gcc.target/gcn/simd-math-5-8.c: New test. * gcc.target/gcn/simd-math-5-char-16.c: New test. * gcc.target/gcn/simd-math-5-char-32.c: New test. * gcc.target/gcn/simd-math-5-char-4.c: New test. * gcc.target/gcn/simd-math-5-char-8.c: New test. * gcc.target/gcn/simd-math-5-char-run-16.c: New test. * gcc.target/gcn/simd-math-5-char-run-32.c: New test. * gcc.target/gcn/simd-math-5-char-run-4.c: New test. * gcc.target/gcn/simd-math-5-char-run-8.c: New test. * gcc.target/gcn/simd-math-5-char-run.c: New test. * gcc.target/gcn/simd-math-5-char.c: New test. * gcc.target/gcn/simd-math-5-long-16.c: New test. * gcc.target/gcn/simd-math-5-long-32.c: New test. * gcc.target/gcn/simd-math-5-long-4.c: New test. * gcc.target/gcn/simd-math-5-long-8.c: New test. * gcc.target/gcn/simd-math-5-long-run-16.c: New test. * gcc.target/gcn/simd-math-5-long-run-32.c: New test. * gcc.target/gcn/simd-math-5-long-run-4.c: New test. * gcc.target/gcn/simd-math-5-long-run-8.c: New test. * gcc.target/gcn/simd-math-5-long-run.c: New test. * gcc.target/gcn/simd-math-5-long.c: New test. * gcc.target/gcn/simd-math-5-run-16.c: New test. * gcc.target/gcn/simd-math-5-run-32.c: New test. * gcc.target/gcn/simd-math-5-run-4.c: New test. * gcc.target/gcn/simd-math-5-run-8.c: New test. * gcc.target/gcn/simd-math-5-run.c: New test. * gcc.target/gcn/simd-math-5-short-16.c: New test. * gcc.target/gcn/simd-math-5-short-32.c: New test. * gcc.target/gcn/simd-math-5-short-4.c: New test. * gcc.target/gcn/simd-math-5-short-8.c: New test. * gcc.target/gcn/simd-math-5-short-run-16.c: New test. * gcc.target/gcn/simd-math-5-short-run-32.c: New test. * gcc.target/gcn/simd-math-5-short-run-4.c: New test. * gcc.target/gcn/simd-math-5-short-run-8.c: New test. * gcc.target/gcn/simd-math-5-short-run.c: New test. * gcc.target/gcn/simd-math-5-short.c: New test. * gcc.target/gcn/simd-math-5.c: New test. (cherry picked from commit d9d6774527bccc5ce0394851aa232f8abdaade4c)
2023-06-20vect: vectorize via libfuncsAndrew Stubbs3-3/+17
This patch allows vectorization when the libfuncs are defined. gcc/ChangeLog: * tree-vect-generic.cc: Include optabs-libfuncs.h. (get_compute_type): Check optab_libfunc. * tree-vect-stmts.cc: Include optabs-libfuncs.h. (vectorizable_operation): Check optab_libfunc. (cherry picked from commit 1f97a0b412ed0984ff4af381b222b87424c08dbd)
2023-06-20amdgcn: minimal V64TImode vector supportAndrew Stubbs4-134/+346
Just enough support for TImode vectors to exist, load, store, move, without any real instructions available. This is primarily for the use of divmodv64di4, which uses TImode to return a pair of DImode values. gcc/ChangeLog: * config/gcn/gcn-protos.h (vgpr_4reg_mode_p): New function. * config/gcn/gcn-valu.md (V_4REG, V_4REG_ALT): New iterators. (V_MOV, V_MOV_ALT): Likewise. (scalar_mode, SCALAR_MODE): Add TImode. (vnsi, VnSI, vndi, VnDI): Likewise. (vec_merge, vec_merge_with_clobber, vec_merge_with_vcc): Use V_MOV. (mov<mode>, mov<mode>_unspec): Use V_MOV. (*mov<mode>_4reg): New insn. (mov<mode>_exec): New 4reg variant. (mov<mode>_sgprbase): Likewise. (reload_in<mode>, reload_out<mode>): Use V_MOV. (vec_set<mode>): Likewise. (vec_duplicate<mode><exec>): New 4reg variant. (vec_extract<mode><scalar_mode>): Likewise. (vec_extract<V_ALL:mode><V_ALL_ALT:mode>): Rename to ... (vec_extract<V_MOV:mode><V_MOV_ALT:mode>): ... this, and use V_MOV. (vec_extract<V_4REG:mode><V_4REG_ALT:mode>_nop): New 4reg variant. (fold_extract_last_<mode>): Use V_MOV. (vec_init<V_ALL:mode><V_ALL_ALT:mode>): Rename to ... (vec_init<V_MOV:mode><V_MOV_ALT:mode>): ... this, and use V_MOV. (gather_load<mode><vnsi>, gather<mode>_expr<exec>, gather<mode>_insn_1offset<exec>, gather<mode>_insn_1offset_ds<exec>, gather<mode>_insn_2offsets<exec>): Use V_MOV. (scatter_store<mode><vnsi>, scatter<mode>_expr<exec_scatter>, scatter<mode>_insn_1offset<exec_scatter>, scatter<mode>_insn_1offset_ds<exec_scatter>, scatter<mode>_insn_2offsets<exec_scatter>): Likewise. (maskload<mode>di, maskstore<mode>di, mask_gather_load<mode><vnsi>, mask_scatter_store<mode><vnsi>): Likewise. * config/gcn/gcn.cc (gcn_class_max_nregs): Use vgpr_4reg_mode_p. (gcn_hard_regno_mode_ok): Likewise. (GEN_VNM): Add TImode support. (USE_TI): New macro. Separate TImode operations from non-TImode ones. (gcn_vector_mode_supported_p): Add V64TImode, V32TImode, V16TImode, V8TImode, and V2TImode. (print_operand): Add 'J' and 'K' print codes. (cherry picked from commit 8aeabd9f63d8a54a5fa0b038ad4425a999e1cc75)
2023-06-19testsuite: Check int128 effective target for pr109932-{1,2}.c [PR110230]Kewen Lin2-0/+2
This patch is to make newly added test cases pr109932-{1,2}.c check int128 effective target to avoid unsupported type error on 32-bit. I did hit this failure during testing and fixed it, but made a stupid mistake not updating the local formatted patch which was actually out of date. PR testsuite/110230 PR target/109932 gcc/testsuite/ChangeLog: * gcc.target/powerpc/pr109932-1.c: Adjust with int128 effective target. * gcc.target/powerpc/pr109932-2.c: Ditto. (cherry picked from commit 16eb9d69079d769b2aa2c07ce54aca20f5547c14)
2023-06-19rs6000: Guard __builtin_{un,}pack_vector_int128 with vsx [PR109932]Kewen Lin3-7/+39
As PR109932 shows, builtins __builtin_{un,}pack_vector_int128 should be guarded under vsx rather than power7, as their corresponding bif patterns have the conditions TARGET_VSX and VECTOR_MEM_ALTIVEC_OR_VSX_P (V1TImode). This patch is to move __builtin_{un,}pack_vector_int128 to stanza vsx to ensure their supports. PR target/109932 gcc/ChangeLog: * config/rs6000/rs6000-builtins.def (__builtin_pack_vector_int128, __builtin_unpack_vector_int128): Move from stanza power7 to vsx. gcc/testsuite/ChangeLog: * gcc.target/powerpc/pr109932-1.c: New test. * gcc.target/powerpc/pr109932-2.c: New test. (cherry picked from commit ff83d1b47aadcdaf80a4fda84b0dc00bb2cd3641)
2023-06-19rs6000: Don't use TFmode for 128 bits fp constant in toc [PR110011]Kewen Lin2-1/+43
As PR110011 shows, when encoding 128 bits fp constant into toc, we adopts REAL_VALUE_TO_TARGET_LONG_DOUBLE which is to find the first float mode with LONG_DOUBLE_TYPE_SIZE bits of precision, it would be TFmode here. But the 128 bits fp constant can be with mode IFmode or KFmode, which doesn't necessarily have the same underlying float format as the one of TFmode, like this PR exposes, with option -mabi=ibmlongdouble TFmode has ibm_extended_format while KFmode has ieee_quad_format, mixing up the formats (the encoding/decoding ways) would cause unexpected results. This patch is to make it use constant's own mode instead of TFmode for real_to_target call. PR target/110011 gcc/ChangeLog: * config/rs6000/rs6000.cc (output_toc): Use the mode of the 128-bit floating constant itself for real_to_target call. gcc/testsuite/ChangeLog: * gcc.target/powerpc/pr110011.c: New test. (cherry picked from commit 388809f2afde874180da0669c669e241037eeba0)
2023-06-20Daily bump.GCC Administrator1-1/+1
2023-06-19OpenACC: Improve implicit mapping for non-lexically nested offload regionsJulian Brown10-18/+31
This patch enables use of the OMP_CLAUSE_RUNTIME_IMPLICIT_P flag for OpenACC. This allows code like this to work correctly: int arr[100]; [...] #pragma acc enter data copyin(arr[20:10]) /* No explicit mapping of 'arr' here. */ #pragma acc parallel { /* use of arr[20:10]... */ } #pragma acc exit data copyout(arr[20:10]) Otherwise, the implicit "copy" ("present_or_copy") on the parallel corresponds to the whole array, and that fails at runtime when the subarray is mapped. The numbering of the GOMP_MAP_IMPLICIT bit clashes with the OpenACC "non-contiguous" dynamic array support, so the GOMP_MAP_NONCONTIG_ARRAY_P macro has been adjusted to account for that. This behaviour relates to upstream OpenACC issue 490 (not yet resolved). 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/ * gimplify.cc (gimplify_adjust_omp_clauses_1): Set OMP_CLAUSE_MAP_RUNTIME_IMPLICIT_P for OpenACC also. gcc/testsuite/ * c-c++-common/goacc/combined-reduction.c: Adjust scan output. * c-c++-common/goacc/reduction-1.c: Likewise. * c-c++-common/goacc/reduction-2.c: Likewise. * c-c++-common/goacc/reduction-3.c: Likewise. * c-c++-common/goacc/reduction-4.c: Likewise. * c-c++-common/goacc/reduction-10.c: Likewise. * gfortran.dg/goacc/loop-tree-1.f90: Likewise. include/ * gomp-constants.h (GOMP_MAP_NONCONTIG_ARRAY_P): Tweak condition. libgomp/ * testsuite/libgomp.oacc-c-c++-common/implicit-mapping-1.c: New test.
2023-06-19OpenACC: Allow implicit uses of assumed-size arrays in offload regionsJulian Brown4-6/+23
This patch reimplements the functionality of the previously-reverted patch "Assumed-size arrays with non-lexical data mappings". The purpose is to support implicit uses of assumed-size arrays for Fortran when those arrays have already been mapped on the target some other way (e.g. by "acc enter data"). This relates to upstream OpenACC issue 489 (not yet resolved). 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/fortran/ * trans-openmp.cc (gfc_omp_finish_clause): Treat implicitly-mapped assumed-size arrays as zero-sized for OpenACC, rather than an error. gcc/testsuite/ * gfortran.dg/goacc/assumed-size.f90: Don't expect error. libgomp/ * testsuite/libgomp.oacc-fortran/nonlexical-assumed-size-1.f90: New test. * testsuite/libgomp.oacc-fortran/nonlexical-assumed-size-2.f90: New test.
2023-06-19OpenACC: "declare create" fixes wrt. "allocatable" variablesJulian Brown4-5/+64
This patch fixes a case revealed by the previous patch where a synthetic "acc data" region created for a "declare create" variable could interact strangely with lexical inheritance behaviour. In fact, it doesn't seem right to create the "acc data" region for allocatable variables at all -- doing so means that a data region is likely to be created for an unallocated variable. The fix is not to add such variables to the synthetic "acc data" region at all, and defer to the code that performs "enter data"/"exit data" for them when allocated/deallocated on the host instead. Then, "declare create" variables are implicitly turned into "present" clauses on in-scope offload regions. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/fortran/ * trans-openmp.cc (gfc_omp_finish_clause): Handle "declare create" for scalar allocatable variables. (gfc_trans_omp_clauses): Don't include allocatable vars in synthetic "acc data" region created for "declare create" variables. Mark such variables with the "oacc declare create" attribute instead. Don't create ALWAYS_POINTER mapping for target-to-host updates of declare create variables. (gfc_trans_oacc_declare): Handle empty clause list. gcc/ * gimplify.cc (gimplify_adjust_omp_clauses_1): Handle "oacc declare create" attribute. libgomp/ * testsuite/libgomp.oacc-fortran/declare-create-1.f90: New test. * testsuite/libgomp.oacc-fortran/declare-create-2.f90: New test. * testsuite/libgomp.oacc-fortran/declare-create-3.f90: New test.
2023-06-19OpenACC: Reimplement "inheritance" for lexically-nested offload regionsJulian Brown8-61/+230
This patch reimplements "lexical inheritance" for OpenACC offload regions inside "data" regions, allowing e.g. this to work: int *ptr; [...] #pragma acc data copyin(ptr[10:2]) { #pragma acc parallel { ... } } here, the "copyin" is mirrored on the inner "acc parallel" as "present(ptr[10:2])" -- allowing code within the parallel to use that section of the array even though the mapping is implicit. In terms of implementation, this works by expanding mapping nodes for "acc data" to include pointer mappings that might be needed by inner offload regions. The resulting mapping group is then copied to the inner offload region as needed, rewriting the first node to "force_present". The pointer mapping nodes are then removed from the "acc data" later during gimplification. For OpenMP, pointer mapping nodes on equivalent "omp data" regions are not needed, so remain suppressed during expansion. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/c-family/ * c-omp.cc (c_omp_address_inspector::expand_array_base): Don't omit pointer nodes for OpenACC. gcc/ * gimplify.cc (omp_tsort_mark, omp_mapping_group): Move before gimplify_omp_ctx. Add constructor to omp_mapping_group. (gimplify_omp_ctx): Add DECL_DATA_CLAUSE field. (new_omp_context, delete_omp_context): Initialise and free above field. (omp_gather_mapping_groups_1): Use constructor for omp_mapping_group. (gimplify_scan_omp_clauses): Record mappings that might be lexically inherited. Don't remove GOMP_MAP_FIRSTPRIVATE_POINTER/GOMP_MAP_FIRSTPRIVATE_REFERENCE yet. (gomp_oacc_needs_data_present): New function. (gimplify_adjust_omp_clauses_1): Implement lexical inheritance behaviour for OpenACC. (gimplify_adjust_omp_clauses): Remove GOMP_MAP_FIRSTPRIVATE_POINTER/GOMP_MAP_FIRSTPRIVATE_REFERENCE here instead, after lexical inheritance is done. gcc/testsuite/ * c-c++-common/goacc/acc-data-chain.c: Re-enable scan test. * gfortran.dg/goacc/pr70828.f90: Likewise. * gfortran.dg/goacc/assumed-size.f90: New test. libgomp/ * testsuite/libgomp.oacc-c-c++-common/pr70828.c: Un-XFAIL. * testsuite/libgomp.oacc-c-c++-common/pr70828-2.c: Un-XFAIL. * testsuite/libgomp.oacc-fortran/pr70828.f90: Un-XFAIL. * testsuite/libgomp.oacc-fortran/pr70828-2.f90: Un-XFAIL. * testsuite/libgomp.oacc-fortran/pr70828-3.f90: Un-XFAIL. * testsuite/libgomp.oacc-fortran/pr70828-4.f90: Un-XFAIL. * testsuite/libgomp.oacc-fortran/pr70828-5.f90: Un-XFAIL. * testsuite/libgomp.oacc-fortran/pr70828-6.f90: Un-XFAIL.
2023-06-19OpenMP/OpenACC: Reorganise OMP map clause handling in gimplify.ccJulian Brown4-291/+389
This patch has been separated out from the C++ "declare mapper" support patch. It contains just the gimplify.cc rearrangement work, mostly moving gimplification from gimplify_scan_omp_clauses to gimplify_adjust_omp_clauses for map clauses. The motivation for doing this was that we don't know if we need to instantiate mappers implicitly until the body of an offload region has been scanned, i.e. in gimplify_adjust_omp_clauses, but we also need the un-gimplified form of clauses to sort by base-pointer dependencies after mapper instantiation has taken place. The patch also reimplements the "present" clause sorting code to avoid another sorting pass on mapping nodes. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/ * gimplify.cc (omp_segregate_mapping_groups): Handle "present" groups. (gimplify_scan_omp_clauses): Use mapping group functionality to iterate through mapping nodes. Remove most gimplification of OMP_CLAUSE_MAP nodes from here, but still populate ctx->variables splay tree. (gimplify_adjust_omp_clauses): Move most gimplification of OMP_CLAUSE_MAP nodes here. gcc/testsuite/ * gfortran.dg/gomp/map-12.f90: Adjust scan output.
2023-06-19OpenMP/OpenACC: Unordered/non-constant component offset runtime diagnosticJulian Brown6-18/+110
This patch adds support for non-constant component offsets in "map" clauses for OpenMP (and the equivalants for OpenACC), which are not able to be sorted into order at compile time. Normally struct accesses in such clauses are gathered together and sorted into increasing address order after a "GOMP_MAP_STRUCT" node: if we have variable indices, that is no longer possible. This version of the patch scales back the previously-posted version to merely add a diagnostic for incorrect usage of component accesses with variably-indexed arrays of structs: the only permitted variant is where we have multiple indices that are the same, but we could not prove so at compile time. Rather than silently producing the wrong result for cases where the indices are in fact different, we error out (e.g., "map(dtarr(i)%arrptr, dtarr(j)%arrptr(4:8))", for different i/j). For now, multiple *constant* array indices are still supported (see map-arrayofstruct-1.c). That could perhaps be addressed with a follow-up patch, if necessary. This version of the patch renumbers the GOMP_MAP_STRUCT_UNORD kind to avoid clashing with the OpenACC "non-contiguous" dynamic array support. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/fortran/ * trans-openmp.cc (gfc_omp_deep_map_kind_p): Add GOMP_MAP_STRUCT_UNORD. gcc/ * gimplify.cc (extract_base_bit_offset): Add VARIABLE_OFFSET parameter. (omp_get_attachment, omp_group_last, omp_group_base, omp_directive_maps_explicitly): Add GOMP_MAP_STRUCT_UNORD support. (omp_accumulate_sibling_list): Update calls to extract_base_bit_offset. Support GOMP_MAP_STRUCT_UNORD. (omp_build_struct_sibling_lists, gimplify_scan_omp_clauses, gimplify_adjust_omp_clauses, gimplify_omp_target_update): Add GOMP_MAP_STRUCT_UNORD support. * omp-low.cc (lower_omp_target): Add GOMP_MAP_STRUCT_UNORD support. * tree-pretty-print.cc (dump_omp_clause): Likewise. include/ * gomp-constants.h (gomp_map_kind): Add GOMP_MAP_STRUCT_UNORD. libgomp/ * oacc-mem.c (find_group_last, goacc_enter_data_internal, goacc_exit_data_internal, GOACC_enter_exit_data): Add GOMP_MAP_STRUCT_UNORD support. * target.c (gomp_map_vars_internal): Add GOMP_MAP_STRUCT_UNORD support. Detect incorrect use of variable indexing of arrays of structs. (GOMP_target_enter_exit_data, gomp_target_task_fn): Add GOMP_MAP_STRUCT_UNORD support. * testsuite/libgomp.c-c++-common/map-arrayofstruct-1.c: New test. * testsuite/libgomp.c-c++-common/map-arrayofstruct-2.c: New test. * testsuite/libgomp.c-c++-common/map-arrayofstruct-3.c: New test. * testsuite/libgomp.fortran/map-subarray-5.f90: New test.
2023-06-19OpenMP: Pointers and member mappingsJulian Brown11-81/+610
This patch changes the mapping node arrangement used for array components of derived types, e.g.: type T integer, pointer, dimension(:) :: arrptr end type T type(T) :: tvar [...] !$omp target map(tofrom: tvar%arrptr) This will currently be mapped using three mapping nodes: GOMP_MAP_TO tvar%arrptr (the descriptor) GOMP_MAP_TOFROM *tvar%arrptr%data (the actual array data) GOMP_MAP_ALWAYS_POINTER tvar%arrptr%data (a pointer to the array data) This follows OMP 5.0, 2.19.7.1 (or OpenMP 5.2, 5.8.3) "map Clause": "If a list item in a map clause is an associated pointer and the pointer is not the base pointer of another list item in a map clause on the same construct, then it is treated as if its pointer target is implicitly mapped in the same clause. For the purposes of the map clause, the mapped pointer target is treated as if its base pointer is the associated pointer." However, we can also write this: map(to: tvar%arrptr) map(tofrom: tvar%arrptr(3:8)) and then instead we should follow (OpenMP 5.2, 5.8.3 "map Clause"): "For map clauses on map-entering constructs, if any list item has a base pointer for which a corresponding pointer exists in the data environment upon entry to the region and either a new list item or the corresponding pointer is created in the device data environment on entry to the region, then: 1. [Fortran] The corresponding pointer variable is associated with a pointer target that has the same rank and bounds as the pointer target of the original pointer, such that the corresponding list item can be accessed through the pointer in a target region. 2. The corresponding pointer variable becomes an attached pointer for the corresponding list item." With this patch you can write the above mappings, and the mapping nodes used to map pointers to array sections (with descriptors) now look like this: 1) map(to: tvar%arrptr) --> GOMP_MAP_TO [implicit] *tvar%arrptr%data (the array data) GOMP_MAP_TO_PSET tvar%arrptr (the descriptor) GOMP_MAP_ATTACH_DETACH tvar%arrptr%data 2) map(tofrom: tvar%arrptr(3:8) --> GOMP_MAP_TOFROM *tvar%arrptr%data(3) (size 8-3+1, etc.) GOMP_MAP_TO_PSET tvar%arrptr GOMP_MAP_ATTACH_DETACH tvar%arrptr%data (bias 3, etc.) In this case, we can determine in the front-end that the whole-array/pointer mapping (1) is only needed to map the pointer -- so we drop it entirely. (Note also that we set -- early -- the OMP_CLAUSE_MAP_RUNTIME_IMPLICIT_P flag for whole-array-via-pointer mappings. See below.) In the middle end, we process mappings using the struct sibling-list handling machinery by moving the "GOMP_MAP_TO_PSET" node from the middle of the group of three mapping nodes to the proper sorted position after the GOMP_MAP_STRUCT mapping: GOMP_MAP_STRUCT tvar (len: 1) GOMP_MAP_TO_PSET tvar%arr (size: 64, etc.) <--. moved here [...] | GOMP_MAP_TOFROM *tvar%arrptr%data(3) ___| GOMP_MAP_ATTACH_DETACH tvar%arrptr%data In another case, if we have an array of derived-type values "dtarr", and mappings like: i = 1 j = 1 map(to: dtarr(i)%arrptr) map(tofrom: dtarr(j)%arrptr(3:8)) We still map the same way, but this time we cannot prove that the base expressions "dtarr(i) and "dtarr(j)" are the same in the front-end. So we keep both mappings, but we move the "[implicit]" mapping of the full-array reference to the end of the clause list in gimplify.cc (by adjusting the topological sorting algorithm): GOMP_MAP_STRUCT dtvar (len: 2) GOMP_MAP_TO_PSET dtvar(i)%arrptr GOMP_MAP_TO_PSET dtvar(j)%arrptr [...] GOMP_MAP_TOFROM *dtvar(j)%arrptr%data(3) (size: 8-3+1) GOMP_MAP_ATTACH_DETACH dtvar(j)%arrptr%data GOMP_MAP_TO [implicit] *dtvar(i)%arrptr%data(1) (size: whole array) GOMP_MAP_ATTACH_DETACH dtvar(i)%arrptr%data Always moving "[implicit]" full-array mappings after array-section mappings (without that bit set) means that we'll avoid copying the whole array unnecessarily -- even in cases where we can't prove that the arrays are the same. The patch also fixes some bugs with "enter data" and "exit data" directives with this new mapping arrangement. Also now if you have mappings like this: #pragma omp target enter data map(to: dv, dv%arr(1:20)) The whole of the derived-type variable "dv" is mapped, so the GOMP_MAP_TO_PSET for the array-section mapping can be dropped: GOMP_MAP_TO dv GOMP_MAP_TO *dv%arr%data GOMP_MAP_TO_PSET dv%arr <-- deleted (array section mapping) GOMP_MAP_ATTACH_DETACH dv%arr%data To accommodate for recent changes to mapping nodes made by Tobias, this version of the patch avoids using GOMP_MAP_TO_PSET for "exit data" directives, in favour of using the "correct" GOMP_MAP_RELEASE/GOMP_MAP_DELETE kinds during early expansion. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/fortran/ * dependency.cc (gfc_omp_expr_prefix_same): New function. * dependency.h (gfc_omp_expr_prefix_same): Add prototype. * gfortran.h (gfc_omp_namelist): Add "duplicate_of" field to "u2" union. * trans-openmp.cc (dependency.h): Include. (gfc_trans_omp_array_section): Adjust mapping node arrangement for array descriptors. Use GOMP_MAP_TO_PSET or GOMP_MAP_RELEASE/GOMP_MAP_DELETE with the OMP_CLAUSE_RELEASE_DESCRIPTOR flag set. (gfc_symbol_rooted_namelist): New function. (gfc_trans_omp_clauses): Check subcomponent and subarray/element accesses elsewhere in the clause list for pointers to derived types or array descriptors, and adjust or drop mapping nodes appropriately. Adjust for changes to mapping node arrangement. (gfc_trans_oacc_executable_directive): Pass code op through. gcc/ * gimplify.cc (omp_map_clause_descriptor_p): New function. (build_omp_struct_comp_nodes, omp_get_attachment, omp_group_base): Use above function. (omp_tsort_mapping_groups): Process nodes that have OMP_CLAUSE_MAP_RUNTIME_IMPLICIT_P set after those that don't. Add enter_exit_data parameter. (omp_resolve_clause_dependencies): Remove GOMP_MAP_TO_PSET mappings if we're mapping the whole containing derived-type variable. (omp_accumulate_sibling_list): Adjust GOMP_MAP_TO_PSET handling. Remove GOMP_MAP_ALWAYS_POINTER handling. (gimplify_scan_omp_clauses): Pass enter_exit argument to omp_tsort_mapping_groups. Don't adjust/remove GOMP_MAP_TO_PSET mappings for derived-type components here. * tree.h (OMP_CLAUSE_RELEASE_DESCRIPTOR): New macro. gcc/testsuite/ * gfortran.dg/gomp/map-9.f90: Adjust scan output. * gfortran.dg/gomp/map-subarray-2.f90: New test. * gfortran.dg/gomp/map-subarray.f90: New test. libgomp/ * testsuite/libgomp.fortran/map-subarray.f90: New test. * testsuite/libgomp.fortran/map-subarray-2.f90: New test. * testsuite/libgomp.fortran/map-subarray-3.f90: New test. * testsuite/libgomp.fortran/map-subarray-4.f90: New test. * testsuite/libgomp.fortran/map-subarray-6.f90: New test. * testsuite/libgomp.fortran/map-subarray-7.f90: New test. * testsuite/libgomp.fortran/map-subarray-8.f90: New test. * testsuite/libgomp.fortran/map-subcomponents.f90: New test. * testsuite/libgomp.fortran/struct-elem-map-1.f90: Adjust for descriptor-mapping changes. Remove XFAIL.
2023-06-19OpenMP: implicitly map base pointer for array-section pointer componentsJulian Brown4-4/+17
Following from discussion in: https://gcc.gnu.org/pipermail/gcc-patches/2021-May/570075.html and: https://gcc.gnu.org/pipermail/gcc-patches/2022-December/608100.html and also upstream OpenMP issue 342, this patch changes mapping for array sections of pointer components on compute regions like this: #pragma omp target map(s.ptr[0:10]) { ...use of 's'... } so the base pointer 's.ptr' is implicitly mapped, and thus pointer attachment happens. This is subtly different in the "enter data" case, e.g: #pragma omp target enter data map(s.ptr[0:10]) if 's.ptr' (or the whole of 's') is not present on the target before the directive is executed, the array section is copied to the target but pointer attachment does *not* take place, since 's' (or 's.ptr') is not mapped implicitly for "enter data". To get a pointer attachment with "enter data", you can do, e.g: #pragma omp target enter data map(s.ptr, s.ptr[0:10]) #pragma omp target { ...implicit use of 's'... } That is, once the attachment has happened, implicit mapping of 's' and uses of 's.ptr[...]' work correctly in the target region. ChangeLog 2022-12-12 Julian Brown <julian@codesourcery.com> gcc/ * gimplify.cc (omp_accumulate_sibling_list): Don't require explicitly-mapped base pointer for compute regions. gcc/testsuite/ * c-c++-comon/gomp/target-implicit-map-2.c: Update expected scan output. libgomp/ * testsuite/libgomp.c-c++-common/target-implicit-map-2.c: Fix missing "free". * testsuite/libgomp.c-c++-common/target-implicit-map-3.c: New test. * testsuite/libgomp.c-c++-common/target-map-zlas-1.c: New test. * testsuite/libgomp.c/target-22.c: Remove explicit base pointer mappings.
2023-06-19OpenMP/OpenACC: Rework clause expansion and nested struct handlingJulian Brown26-813/+2909
This patch is an extension and rewrite/rethink of the following two patches: "OpenMP/OpenACC: Add inspector class to unify mapped address analysis" https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591977.html "OpenMP: Handle reference-typed struct members" https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591978.html The latter was reviewed here by Jakub: https://gcc.gnu.org/pipermail/gcc-patches/2022-May/595510.html with the with the comment, > Why isn't a reference to pointer handled that way too? and that opened a whole can of worms... generally, C++ references were not handled very consistently after the clause-processing code had been extended several times already for both OpenACC and OpenMP, and many cases of using C++ (and Fortran) references were broken. Even some cases not involving references were being mapped incorrectly. At present a single clause may be turned into several mapping nodes, or have its mapping type changed, in several places scattered through the front- and middle-end. The analysis relating to which particular transformations are needed for some given expression has become quite hard to follow. Briefly, we manipulate clause types in the following places: 1. During parsing, in c_omp_adjust_map_clauses. Depending on a set of rules, we may change a FIRSTPRIVATE_POINTER (etc.) mapping into ATTACH_DETACH, or mark the decl addressable. 2. In semantics.cc or c-typeck.cc, clauses are expanded in handle_omp_array_sections (called via {c_}finish_omp_clauses, or in finish_omp_clauses itself. The two cases are for processing array sections (the former), or non-array sections (the latter). 3. In gimplify.cc, we build sibling lists for struct accesses, which groups and sorts accesses along with their struct base, creating new ALLOC/RELEASE nodes for pointers. 4. In gimplify.cc:gimplify_adjust_omp_clauses, mapping nodes may be adjusted or created. This patch doesn't completely disrupt this scheme, though clause types are no longer adjusted in c_omp_adjust_map_clauses (step 1). Clause expansion in step 2 (for C and C++) now uses a single, unified mechanism, parts of which are also reused for analysis in step 3. Rather than the kind-of "ad-hoc" pattern matching on addresses used to expand clauses used at present, a new method for analysing addresses is introduced. This does a recursive-descent tree walk on expression nodes, and emits a vector of tokens describing each "part" of the address. This tokenized address can then be translated directly into mapping nodes, with the assurance that no part of the expression has been inadvertently skipped or misinterpreted. In this way, all the variations of ways pointers, arrays, references and component accesses might be combined can be teased apart into easily-understood cases - and we know we've "parsed" the whole address before we start analysis, so the right code paths can easily be selected. For example, a simple access "arr[idx]" might parse as: base-decl access-indexed-array or "mystruct->foo[x]" with a pointer "foo" component might parse as: base-decl access-pointer component-selector access-pointer A key observation is that support for "array" bases, e.g. accesses whose root nodes are not structures, but describe scalars or arrays, and also *one-level deep* structure accesses, have first-class support in gimplify and beyond. Expressions that use deeper struct accesses or e.g. multiple indirections were more problematic: some cases worked, but lots of cases didn't. This patch reimplements the support for those in gimplify.cc, again using the new "address tokenization" support. An expression like "mystruct->foo->bar[0:10]" used in a mapping node will translate the right-hand access directly in the front-end. The base for the access will be "mystruct->foo". This is handled recursively in gimplify.cc -- there may be several accesses of "mystruct"'s members on the same directive, so the sibling-list building machinery can be used again. (This was already being done for OpenACC, but the new implementation differs somewhat in details, and is more robust.) For OpenMP, in the case where the base pointer itself, i.e. "mystruct->foo" here, is NOT mapped on the same directive, we create a "fragile" mapping. This turns the "foo" component access into a zero-length allocation (which is a new feature for the runtime, so support has been added there too). A couple of changes have been made to how mapping clauses are turned into mapping nodes: The first change is based on the observation that it is probably never correct to use GOMP_MAP_ALWAYS_POINTER for component accesses (e.g. for references), because if the containing struct is already mapped on the target then the host version of the pointer in question will be corrupted if the struct is copied back from the target. This patch removes all such uses, across each of C, C++ and Fortran. The second change is to the way that GOMP_MAP_ATTACH_DETACH nodes are processed during sibling-list creation. For OpenMP, for pointer components, we must map the base pointer separately from an array section that uses the base pointer, so e.g. we must have both "map(mystruct.base)" and "map(mystruct.base[0:10])" mappings. These create nodes such as: GOMP_MAP_TOFROM mystruct.base G_M_TOFROM *mystruct.base [len: 10*elemsize] G_M_ATTACH_DETACH mystruct.base Instead of using the first of these directly when building the struct sibling list then skipping the group using GOMP_MAP_ATTACH_DETACH, leading to: GOMP_MAP_STRUCT mystruct [len: 1] GOMP_MAP_TOFROM mystruct.base we now introduce a new "mini-pass", omp_resolve_clause_dependencies, that drops the GOMP_MAP_TOFROM for the base pointer, marks the second group as having had a base-pointer mapping, then omp_build_struct_sibling_lists can create: GOMP_MAP_STRUCT mystruct [len: 1] GOMP_MAP_ALLOC mystruct.base [len: ptrsize] This ends up working better in many cases, particularly those involving references. (The "alloc" space is immediately overwritten by a pointer attachment, so this is mildly more efficient than a redundant TO mapping at runtime also.) There is support in the address tokenizer for "arbitrary" base expressions which aren't rooted at a decl, but that is not used as present because such addresses are disallowed at parse time. In the front-ends, the address tokenization machinery is mostly only used for clause expansion and not for diagnostics at present. It could be used for those too, which would allow more of my previous "address inspector" implementation to be removed. The new bits in gimplify.cc work with OpenACC also. This version of the patch has been rebased for og13 and incorporates a couple of follow-up patches. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/c-family/ * c-common.h (c_omp_region_type): Add C_ORT_ACC_TARGET. (omp_addr_token): Add forward declaration. (c_omp_address_inspector): New class. * c-omp.cc (c_omp_adjust_map_clauses): Mark decls addressable here, but do not change any mapping node types. (c_omp_address_inspector::unconverted_ref_origin, c_omp_address_inspector::component_access_p, c_omp_address_inspector::check_clause, c_omp_address_inspector::get_root_term, c_omp_address_inspector::map_supported_p, c_omp_address_inspector::get_origin, c_omp_address_inspector::maybe_unconvert_ref, c_omp_address_inspector::maybe_zero_length_array_section, c_omp_address_inspector::expand_array_base, c_omp_address_inspector::expand_component_selector, c_omp_address_inspector::expand_map_clause): New methods. (omp_expand_access_chain): New function. gcc/c/ * c-parser.cc (c_parser_oacc_all_clauses): Add TARGET parameter. Use to select region type for c_finish_omp_clauses call. (c_parser_oacc_loop): Update calls to c_parser_oacc_all_clauses. (c_parser_oacc_compute): Likewise. (c_parser_omp_target_enter_data): Support ATTACH kind. (c_parser_omp_target_exit_data): Support DETACH kind. * c-typeck.cc (handle_omp_array_sections_1, handle_omp_array_sections, c_finish_omp_clauses): Use c_omp_address_inspector class and OMP address tokenizer to analyze and expand map clause expressions. Fix some diagnostics. Fix for C_ORT_ACC_TARGET. gcc/cp/ * parser.cc (cp_parser_oacc_all_clauses): Add TARGET parameter. Use to select region type for finish_omp_clauses call. (cp_parser_omp_target_enter_data): Support GOMP_MAP_ATTACH kind. (cp_parser_omp_target_exit_data): Support GOMP_MAP_DETACH kind. (cp_parser_oacc_declare): Update call to cp_parser_oacc_all_clauses. (cp_parser_oacc_loop): Update calls to cp_parser_oacc_all_clauses. (cp_parser_oacc_compute): Likewise. * pt.cc (tsubst_expr): Use C_ORT_ACC_TARGET for call to tsubst_omp_clauses for compute regions. * semantics.cc (cp_omp_address_inspector): New class, derived from c_omp_address_inspector. (handle_omp_array_sections_1, handle_omp_array_sections, finish_omp_clauses): Use cp_omp_address_inspector class and OMP address tokenizer to analyze and expand OpenMP map clause expressions. Fix some diagnostics. Support C_ORT_ACC_TARGET. gcc/fortran/ * trans-openmp.cc (gfc_trans_omp_array_section): Add OPENMP parameter. Use GOMP_MAP_ATTACH_DETACH instead of GOMP_MAP_ALWAYS_POINTER for derived type components. (gfc_trans_omp_clauses): Update calls to gfc_trans_omp_array_section. gcc/ * gimplify.cc (build_struct_comp_nodes): Don't process GOMP_MAP_ATTACH_DETACH "middle" nodes here. (omp_mapping_group): Add REPROCESS_STRUCT and FRAGILE booleans for nested struct handling. (omp_strip_components_and_deref, omp_strip_indirections): Remove functions. (omp_gather_mapping_groups_1): Initialise reprocess_struct and fragile fields. (omp_group_base): Handle GOMP_MAP_ATTACH_DETACH after GOMP_MAP_STRUCT. (omp_index_mapping_groups_1): Skip reprocess_struct groups. (omp_get_nonfirstprivate_group, omp_directive_maps_explicitly, omp_resolve_clause_dependencies, omp_expand_access_chain, omp_first_chained_access_token): New functions. (omp_check_mapping_compatibility): Adjust accepted node combinations for changes elsewhere. (omp_accumulate_sibling_list): Add GROUP_MAP, ADDR_TOKENS, FRAGILE_P, REPROCESSING_STRUCT, ADDED_TAIL parameters. Use OMP address tokenizer to analyze addresses. Reimplement nested struct handling, and implement "fragile groups". (omp_build_struct_sibling_lists): Adjust for changes to omp_accumulate_sibling_list. Recalculate bias for ATTACH_DETACH nodes after GOMP_MAP_STRUCT nodes. (gimplify_scan_omp_clauses): Call omp_resolve_clause_dependencies. Use OMP address tokenizer. (gimplify_adjust_omp_clauses_1): Use build_fold_indirect_ref_loc instead of build_simple_mem_ref_loc. * omp-general.cc (omp-general.h): Include. (omp_addr_tokenizer): New namespace. (omp_addr_tokenizer::omp_addr_token): New. (omp_addr_tokenizer::omp_parse_component_selector, omp_addr_tokenizer::omp_parse_ref, omp_addr_tokenizer::omp_parse_pointer, omp_addr_tokenizer::omp_parse_access_method, omp_addr_tokenizer::omp_parse_access_methods, omp_addr_tokenizer::omp_parse_structure_base, omp_addr_tokenizer::omp_parse_structured_expr, omp_addr_tokenizer::omp_parse_array_expr, omp_addr_tokenizer::omp_access_chain_p, omp_addr_tokenizer::omp_accessed_addr): New functions. (omp_parse_expr, debug_omp_tokenized_addr): New functions. * omp-general.h (omp_addr_tokenizer::access_method_kinds, omp_addr_tokenizer::structure_base_kinds, omp_addr_tokenizer::token_type, omp_addr_tokenizer::omp_addr_token, omp_addr_tokenizer::omp_access_chain_p, omp_addr_tokenizer::omp_accessed_addr): New. (omp_addr_token, omp_parse_expr): New. * omp-low.cc (scan_sharing_clauses): Skip error check for references to pointers. * tree.h (OMP_CLAUSE_ATTACHMENT_MAPPING_ERASED): New macro. gcc/testsuite/ * c-c++-common/gomp/clauses-2.c: Fix error output. * c-c++-common/gomp/target-implicit-map-2.c: Adjust scan output. * c-c++-common/gomp/target-50.c: Adjust scan output. * c-c++-common/gomp/target-enter-data-1.c: Adjust scan output. * g++.dg/gomp/static-component-1.C: New test. * gcc.dg/gomp/target-3.c: Adjust scan output. * gfortran.dg/gomp/map-9.c.f90: Adjust scan output. libgomp/ * target.c (gomp_map_fields_existing): Use gomp_map_0len_lookup. (gomp_attach_pointer): Allow attaching null pointers (or Fortran "unassociated" pointers). (gomp_map_vars_internal): Handle zero-sized struct members. Add diagnostic for unmapped struct pointer members. * testsuite/libgomp.c++/class-array-1.C: New test. * testsuite/libgomp.c-c++-common/baseptrs-1.c: New test. * testsuite/libgomp.c-c++-common/baseptrs-2.c: New test. * testsuite/libgomp.c-c++-common/ptr-attach-2.c: New test. * testsuite/libgomp.c++/baseptrs-3.C: New test. * testsuite/libgomp.c++/baseptrs-4.C: New test. * testsuite/libgomp.c++/baseptrs-5.C: New test. * testsuite/libgomp.c++/target-48.C: New test. * testsuite/libgomp.c++/target-49.C: New test. * testsuite/libgomp.c/target-22.c: Add necessary explicit base pointer mappings. * testsuite/libgomp.fortran/struct-elem-map-1.f90: Add XFAIL.
2023-06-19OpenMP/OpenACC: Reindent TO/FROM/_CACHE_ stanza in {c_}finish_omp_clauseJulian Brown4-682/+703
This patch trivially adds braces and reindents the OMP_CLAUSE_TO/OMP_CLAUSE_FROM/OMP_CLAUSE__CACHE_ stanza in c_finish_omp_clause and finish_omp_clause, in preparation for the following patch (to clarify the diff a little). 2022-09-13 Julian Brown <julian@codesourcery.com> gcc/c/ * c-typeck.cc (c_finish_omp_clauses): Add braces and reindent OMP_CLAUSE_TO/OMP_CLAUSE_FROM/OMP_CLAUSE__CACHE_ stanza. gcc/cp/ * semantics.cc (finish_omp_clause): Add braces and reindent OMP_CLAUSE_TO/OMP_CLAUSE_FROM/OMP_CLAUSE__CACHE_ stanza.
2023-06-19Revert "openmp: Handle C/C++ array reference base-pointers in array sections"Julian Brown6-55/+26
This reverts commit 3385743fd2fa15a2a750a29daf6d4f97f5aad0ae. 2023-06-16 Julian Brown <julian@codesourcery.com> Revert: 2022-02-24 Chung-Lin Tang <cltang@codesourcery.com> gcc/c/ChangeLog: * c-typeck.cc (handle_omp_array_sections): Add handling for creating array-reference base-pointer attachment clause. gcc/cp/ChangeLog: * semantics.cc (handle_omp_array_sections): Add handling for creating array-reference base-pointer attachment clause. gcc/testsuite/ChangeLog: * c-c++-common/gomp/target-enter-data-1.c: Adjust testcase. libgomp/ChangeLog: * testsuite/libgomp.c-c++-common/ptr-attach-2.c: New test.
2023-06-19Revert "Fix implicit mapping for array slices on lexically-enclosing data ↵Julian Brown8-158/+40
constructs (PR70828)" This reverts commit a84b89b8f070f1efe86ea347e98d57e6bc32ae2d. Relevant tests are temporarily disabled or XFAILed. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/ Revert: * gimplify.cc (oacc_array_mapping_info): New struct. (gimplify_omp_ctx): Add decl_data_clause hash map. (new_omp_context): Zero-initialise above. (delete_omp_context): Delete above if allocated. (gimplify_scan_omp_clauses): Scan for array mappings on data constructs, and record in above map. (gomp_oacc_needs_data_present): New function. (gimplify_adjust_omp_clauses_1): Handle data mappings (e.g. array slices) declared in lexically-enclosing data constructs. * omp-low.cc (lower_omp_target): Allow decl for bias not to be present in OpenACC context. gcc/fortran/ Revert: * trans-openmp.cc: Handle implicit "present". gcc/testsuite/ * c-c++-common/goacc/acc-data-chain.c: Partly disable test. * gfortran.dg/goacc/pr70828.f90: Likewise. libgomp/ * testsuite/libgomp.oacc-c-c++-common/pr70828.c: XFAIL test. * testsuite/libgomp.oacc-c-c++-common/pr70828-2.c: XFAIL test. * testsuite/libgomp.oacc-fortran/pr70828.f90: XFAIL test. * testsuite/libgomp.oacc-fortran/pr70828-2.f90: XFAIL test. * testsuite/libgomp.oacc-fortran/pr70828-3.f90: XFAIL test. * testsuite/libgomp.oacc-fortran/pr70828-4.f90: XFAIL test. * testsuite/libgomp.oacc-fortran/pr70828-5.f90: XFAIL test. * testsuite/libgomp.oacc-fortran/pr70828-6.f90: XFAIL test.
2023-06-19Revert "Fix references declared in lexically-enclosing OpenACC data region"Julian Brown4-49/+29
This reverts commit c9cd2bac6a5127a01c6f47e5636a926ac39b5e21. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/fortran/ Revert: * trans-openmp.cc (gfc_omp_finish_clause): Guard addition of clauses for pointers with DECL_P. gcc/ Revert: * gimplify.cc (oacc_array_mapping_info): Add REF field. (gimplify_scan_omp_clauses): Initialise above field for data blocks passed by reference. (gomp_oacc_needs_data_present): Handle references. (gimplify_adjust_omp_clauses_1): Handle references and optional arguments for variables declared in lexically-enclosing OpenACC data region.
2023-06-19Revert "Assumed-size arrays with non-lexical data mappings"Julian Brown5-32/+28
This reverts commit 72733f6e6f6ec1bb9884fea8bfbebd3de03d9374. 2023-06-16 Julian Brown <julian@codesourcery.com> gcc/ Revert: * gimplify.cc (gimplify_adjust_omp_clauses_1): Raise error for assumed-size arrays in map clauses for Fortran/OpenMP. * omp-low.cc (lower_omp_target): Set the size of assumed-size Fortran arrays to one to allow use of data already mapped on the offload device. gcc/fortran/ Revert: * trans-openmp.cc (gfc_omp_finish_clause): Change clauses mapping assumed-size arrays to use the GOMP_MAP_FORCE_PRESENT map type.
2023-06-19Daily bump.GCC Administrator1-1/+1
2023-06-18Daily bump.GCC Administrator1-1/+1
2023-06-17Daily bump.GCC Administrator3-1/+10
2023-06-16RA: Constrain class of pic offset table pseudo to general regsVladimir N. Makarov2-0/+33
On some targets an integer pseudo can be assigned to a FP reg. For pic offset table pseudo it means we will reload the pseudo in this case and, as a consequence, memory containing the pseudo might be recognized as wrong one. The patch fix this problem. PR target/109541 gcc/ChangeLog: * ira-costs.cc: (find_costs_and_classes): Constrain classes of pic offset table pseudo to a general reg subset. gcc/testsuite/ChangeLog: * gcc.target/sparc/pr109541.c: New.
2023-06-16Daily bump.GCC Administrator3-1/+20
2023-06-15ada: Fix internal error on loop iterator filter with -gnatVaEric Botcazou3-4/+13
The problem is that the condition of the iterator filter is expanded early, before it is integrated into an if statement of the loop body, so there is no place to attach the actions generated by this expansion. This happens only for simple loops, i.e. with a parameter specification, so the fix uses the same approach for them as for loops based on iterators. gcc/ada/ * sinfo.ads (Iterator_Filter): Document field. * sem_ch5.adb (Analyze_Iterator_Specification): Move comment around. (Analyze_Loop_Parameter_Specification): Only preanalyze the iterator filter, if any. * exp_ch5.adb (Expand_N_Loop_Statement): Analyze the new list built when an iterator filter is present.
2023-06-15LoongArch: Avoid non-returning indirect jumps through $ra [PR110136]Lulu Cheng1-2/+6
Micro-architecture unconditionally treats a "jr $ra" as "return from subroutine", hence doing "jr $ra" would interfere with both subroutine return prediction and the more general indirect branch prediction. Therefore, a problem like PR110136 can cause a significant increase in branch error prediction rate and affect performance. The same problem exists with "indirect_jump". gcc/ChangeLog: PR target/110136 * config/loongarch/loongarch.md: Modify the register constraints for template "jumptable" and "indirect_jump" from "r" to "e". Co-authored-by: Andrew Pinski <apinski@marvell.com> (cherry picked from commit 5430c86e71927492399129f3df80824c6c334ddf)
2023-06-15Daily bump.GCC Administrator3-1/+20
2023-06-14OpenMP: Fortran support for imperfectly nested loopsSandra Loosemore13-116/+857
OpenMP 5.0 removed the restriction that multiple collapsed loops must be perfectly nested, allowing "intervening code" (including nested BLOCKs) before or after each nested loop. In GCC this code is moved into the inner loop body by the respective front ends. In the Fortran front end, most of the semantic processing happens during the translation phase, so the parse phase just collects the intervening statements, checks them for errors, and splices them around the loop body. gcc/fortran/ChangeLog * openmp.cc: Include omp-api.h. (resolve_omp_clauses): Consolidate inscan reduction clause conflict checking here. (scan_for_next_loop_in_chain): New. (scan_for_next_loop_in_block): New. (gfc_resolve_omp_do_blocks): Set omp_current_do_collapse properly. Handle imperfectly-nested loops when looking for nested omp scan. Refactor to move inscan reduction clause conflict checking to resolve_omp_clauses. (gfc_resolve_do_iterator): Handle imperfectly-nested loops. (struct icode_error_state): New. (icode_code_error_callback): New. (icode_expr_error_callback): New. (diagnose_intervening_code_errors_1): New. (diagnose_intervening_code_errors): New. (restructure_intervening_code): New. (resolve_nested_loops): Update error handling, and extend to detect imperfect nesting errors and check validity of intervening code. Call restructure_intervening_code if needed. (resolve_omp_do): Rename collapse -> count. gcc/testsuite/ChangeLog * gfortran.dg/gomp/collapse1.f90: Adjust expected errors. * gfortran.dg/gomp/collapse2.f90: Likewise. * gfortran.dg/gomp/imperfect1.f90: New. * gfortran.dg/gomp/imperfect2.f90: New. * gfortran.dg/gomp/imperfect3.f90: New. * gfortran.dg/gomp/imperfect4.f90: New. * gfortran.dg/gomp/imperfect5.f90: New. * gfortran.dg/gomp/loop-transforms/tile-1.f90: Adjust expected errors. * gfortran.dg/gomp/loop-transforms/tile-2.f90: Likewise. * gfortran.dg/gomp/loop-transforms/tile-imperfect-nest.f90: Likewise. libgomp/ChangeLog * testsuite/libgomp.fortran/imperfect-destructor.f90: New. * testsuite/libgomp.fortran/imperfect-transform-1.f90: New. * testsuite/libgomp.fortran/imperfect-transform-2.f90: New. * testsuite/libgomp.fortran/imperfect1.f90: New. * testsuite/libgomp.fortran/imperfect2.f90: New. * testsuite/libgomp.fortran/imperfect3.f90: New. * testsuite/libgomp.fortran/imperfect4.f90: New. * testsuite/libgomp.fortran/target-imperfect-transform-1.f90: New. * testsuite/libgomp.fortran/target-imperfect-transform-2.f90: New. * testsuite/libgomp.fortran/target-imperfect1.f90: New. * testsuite/libgomp.fortran/target-imperfect2.f90: New. * testsuite/libgomp.fortran/target-imperfect3.f90: New. * testsuite/libgomp.fortran/target-imperfect4.f90: New.
2023-06-14OpenMP: Refactor and tidy Fortran front-end code for loop transformationsSandra Loosemore6-267/+204
This patch rearranges some code previously added to support loop transformations to simplify merging support for imperfectly-nested loops in a subsequent patch. There is no new functionality added here. gcc/fortran/ChangeLog * openmp.cc (find_nested_loop_in_chain): Move up in file. (find_nested_loop_in_block): Likewise. (resolve_nested_loops): New helper function to consolidate code from... (resolve_omp_do, resolve_omp_tile): ...these functions. Also, remove the redundant call to resolve_nested_loop_transforms, and use uniform error message wording. gcc/testsuite/ChangeLog * gfortran.dg/gomp/collapse1.f90: Adjust expected error message. * gfortran.dg/gomp/collapse2.f90: Likewise. * gfortran.dg/gomp/loop-transforms/tile-2.f90: Likewise.
2023-06-14OpenMP: New c/c++ testcases for imperfectly-nested loopsSandra Loosemore6-0/+203
gcc/testsuite/ChangeLog * c-c++-common/gomp/imperfect1.c: New. * c-c++-common/gomp/imperfect2.c: New. * c-c++-common/gomp/imperfect3.c: New. * c-c++-common/gomp/imperfect4.c: New. * c-c++-common/gomp/imperfect5.c: New. libgomp/ChangeLog * testsuite/libgomp.c-c++-common/imperfect-transform-1.c: New. * testsuite/libgomp.c-c++-common/imperfect-transform-2.c: New. * testsuite/libgomp.c-c++-common/imperfect1.c: New. * testsuite/libgomp.c-c++-common/imperfect2.c: New. * testsuite/libgomp.c-c++-common/imperfect3.c: New. * testsuite/libgomp.c-c++-common/imperfect4.c: New. * testsuite/libgomp.c-c++-common/imperfect5.c: New. * testsuite/libgomp.c-c++-common/imperfect6.c: New. * testsuite/libgomp.c-c++-common/target-imperfect-transform-1.c: New. * testsuite/libgomp.c-c++-common/target-imperfect-transform-2.c: New. * testsuite/libgomp.c-c++-common/target-imperfect1.c: New. * testsuite/libgomp.c-c++-common/target-imperfect2.c: New. * testsuite/libgomp.c-c++-common/target-imperfect3.c: New. * testsuite/libgomp.c-c++-common/target-imperfect4.c: New.
generated by cgit v1.1 at 2025-09-22 15:03:57 +0000