aboutsummaryrefslogtreecommitdiff
path: root/gcc
AgeCommit message (Collapse)AuthorFilesLines
2022-06-30Fortran: fix simplification of INDEX(str1,str2) [PR105691]Harald Anlauf2-102/+60
gcc/fortran/ChangeLog: PR fortran/105691 * simplify.cc (gfc_simplify_index): Replace old simplification code by the equivalent of the runtime library implementation. Use HOST_WIDE_INT instead of int for string index, length variables. gcc/testsuite/ChangeLog: PR fortran/105691 * gfortran.dg/index_6.f90: New test. (cherry picked from commit ff35dbc02092fbcd3d814fcd9fe8e871c3f741fd)
2022-06-30Fortran: fix checking of arguments to UNPACK when MASK is a variable [PR105813]Harald Anlauf2-0/+14
gcc/fortran/ChangeLog: PR fortran/105813 * check.cc (gfc_check_unpack): Try to simplify MASK argument to UNPACK so that checking of the VECTOR argument can work when MASK is a variable. gcc/testsuite/ChangeLog: PR fortran/105813 * gfortran.dg/unpack_vector_1.f90: New test. (cherry picked from commit f21f17f95c0237f4f987a5fa9f1fa9c7e0db3c40)
2022-06-30Fix mis-merge of 'dwarf: Multi-register CFI address support'Kwok Cheung Yeung2-1/+5
This should be a fixup to 13b6c7639cfdca892a3f02b63596b097e1839f38: 'dwarf: Multi-register CFI address support'. 2022-06-30 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * dwarf2cfi.cc (get_cfa_from_loc_descr): Check op against DW_OP_bregx.
2022-06-30Build fixes for OG12 on more recent GCC versionsKwok Cheung Yeung7-8/+32
This fixes a number of minor issues that can cause the build to fail on recent versions of GCC. 2022-06-30 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/fortran/ * openmp.cc (gfc_resolve_omp_allocate): Initialize tail to NULL. (This should be a fixup to 491478d12b83e102f72858e8a871a25c951df293: 'Add parsing support for allocate directive (OpenMP 5.0)') 2022-06-30 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * omp-offload.cc (oacc_loop_get_cfg_loop): Cast tail_mark to gimple* for dump_printf. * tree-scalar-evolution.cc (oacc_ifn_call_extract): Remove unused variable 'call'. (This should be a fixup to 491478d12b83e102f72858e8a871a25c951df293: 'Build fix for 'openacc: Use Graphite for dependence analysis in "kernels" region') 2022-06-30 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * omp-low.cc (usm_transform): Remove unused function argument. (This should be a fixup to 54c2d861ac62e30ebf34a4e62ee0d55478a742b9: 'Build fix for 'openmp: Use libgomp memory allocation functions with unified shared memory') 2022-06-30 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * omp-offload.cc (oacc_loop_warn_if_false_independent): Remove extra '.' at end of message. (This should be a fixup to 01e6774b725ffa667efd818a46795189a281addf: 'Build fix for 'openacc: Warn about "independent" "kernels" loops with data-dependences') 2022-06-30 Tobias Burnus <tobias@codesourcery.com> gcc/ * graphite-isl-ast-to-gimple.cc (graphite_oacc_analyze_scop): Update arguments of dump_printf. (This should be a fixup to a7e863fc4d54fb645fef05f01a024250184964bb: 'openacc: Add runtime alias checking for OpenACC kernels')
2022-06-30Fix string formatting issuesKwok Cheung Yeung3-1/+41
Stricter format-string checking in more recent versions of GCC can cause build failures. 2022-06-30 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * omp-data-optimize.cc (omp_data_optimize_add_candidate): Suppress format checking. (omp_data_optimize_can_be_private): Likewise. (omp_data_optimize_can_be_private): Likewise. (This should be a fixup to ab53d5a6a27dce2a92f28a62ceb6e184c8356f25: 'openacc: Add data optimization pass') 2022-06-30 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * gimplify.cc (gimplify_scan_omp_clauses): Remove extra '%<..%>' pair in format string. (This should be a fixup to dbc770c4351c8824e8083f8aff6117a6b4ba3c0d: 'openmp: Implement uses_allocators clause')
2022-06-30Add ChangeLog.omp entries for "Build fix for 'openmp: allow requires ↵Kwok Cheung Yeung2-0/+8
unified_shared_memory'"
2022-06-30Build fix for 'openmp: allow requires unified_shared_memory'Tobias Burnus2-8/+8
OG12 commit fa65fc45972d27f2fd79a44eaba1978348177ee9 added an error diagnostic (moved around in later commits); this diagnostic caused bootstrap fails as %<...%> were missing. This commit adds them. gcc/c/ * c-parser.cc (c_parser_omp_requires): Add missing %<...%> in error. gcc/cp/ * parser.cc (cp_parser_omp_requires): Add missing %<...%> in error.
2022-06-30Daily bump.GCC Administrator2-1/+19
2022-06-29bootstrap/105551 - restore nvptx buildRichard Biener1-4/+8
The following makes sure to disable var-tracking if only dwarf2-line debuginfo is present. 2022-05-11 Richard Biener <rguenther@suse.de> PR bootstrap/105551 * opts.cc (finish_options): Also disable var-tracking if !DWARF2_DEBUGGING_INFO. (cherry picked from commit e7d9fdf5e0ee4c34a880139254340b4165016289)
2022-06-29LoongArch: Remove undefined behavior from code [PR 106097]Lulu Cheng1-5/+6
C++2017 and previous standard description: The value of E1 << E2 is E1 left-shifted E2 bit positions; vacated bits are zero-filled. If E1 has an unsigned type, the value of the result is E1×2E2, reduced modulo one more than the maximum value representable inthe result type. Otherwise, if E1 has a signed type and non-negative value, and E1×2E2 is representablein the corresponding unsigned type of the result type, then that value, converted to the result type, is the resulting value; otherwise, the behavior is undefined. The value of E1 >> E2 is E1 right-shifted E2 bit positions. If E1 has an unsigned type or if E1 has a signed type and a non-negative value, the value of the result is the integral part of the quotient of E1/2E2. If E1 has a signed type and a negative value, the resulting value is implementation-defined. gcc/ChangeLog: PR target/106097 * config/loongarch/loongarch.cc (loongarch_build_integer): Remove undefined behavior from code. (cherry picked from commit 43653547e7c8da2cd861bceb4a3e4bd338787ced)
2022-06-29Daily bump.GCC Administrator3-1/+55
2022-06-28Fix Fortran array-access regressionsKwok Cheung Yeung2-4/+9
The arguments to gfc_build_array_ref were recently updated in the commit 'fortran: Use pointer arithmetic to index arrays [PR102043]', but a call from gfc_conv_array_ref used the old function signature. This went unnoticed due to the use of default arguments. This patch should be merged into 'Fortran: delinearize multi-dimensional array accesses'. 2022-05-22 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/fortran/ * trans-array.cc (gfc_conv_array_ref): Update arguments to gfc_build_array_ref.
2022-06-28Fix ICE in nested-function-1.c testcaseKwok Cheung Yeung2-1/+14
The ICE occurs during Gimple verification after the ompexp stage because one of the arguments to the generated builtin call is of a Gimple reg type, but isn't a Gimple value (because it is marked addressable). This appears to be fallout from the commit "OpenACC 'kernels' decomposition: Mark variables used in synthesized data clauses as addressable [PR100280]". The launch dimensions have been added to the arguments of a builtin call by oacc_set_fn_attrib, but one of the dimensions has been marked addressable. Fixed by forcing the added arguments to be re-gimplified. 2022-05-13 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * omp-expand.cc (expand_omp_target): Gimplify launch dimensions used in function call.
2022-06-28Fix ICE in kernels-decompose-pr100280-1.c testcaseKwok Cheung Yeung2-0/+7
Check that there is a DECL_INITIAL associated with prev_stmt before using it. 2022-04-15 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/c-family/ * c-omp.cc (check_and_annotate_for_loop): Check that the DECL_INITIAL is non-NULL before using.
2022-06-28Fix ICE in OpenACC kernel testcasesKwok Cheung Yeung2-2/+8
This is due to sequences like this occurring: <bb 11> : .data_dep.8_27 = .UNIQUE (OACC_TAIL_MARK, .data_dep.8_16, 2); <bb 12> : .data_dep.8_29 = .UNIQUE (OACC_JOIN, .data_dep.8_27, -1); ... <bb 15> : .UNIQUE (OACC_TAIL_MARK, .data_dep.8_33); The final tail mark has no LHS, causing code that assumes its presence to segfault. The LHS and the assignment appear to have been removed as dead code by the cddce1 stage. Fixed by checking for the presence of the LHS before using it. 2022-04-14 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * graphite-oacc.cc (find_oacc_tail_marks): Check that data_dep is non-NULL before testing it. (reduction_use_in_outer_loop_p): Likewise.
2022-06-28Fix ICE when cache-3-1.c testcase is runKwok Cheung Yeung4-2/+14
A change that was present in the OG11 version of 'openmp: in_reduction clause support on target construct' but not in the mainline version resulted in non-contiguous arrays being accepted in cache clauses, only to ICE later. 2022-03-17 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/c/ * c-typeck.cc (handle_omp_array_sections_1): Add check to ensure that clause is a map. gcc/cp/ * semantics.cc (handle_omp_array_sections_1): Add check to ensure that clause is a map.
2022-06-28openmp: Add omp_all_memory support (C/C++ only so far)Jakub Jelinek20-35/+381
The ugly part is that OpenMP 5.1 made omp_all_memory a reserved identifier which isn't allowed to be used anywhere but in the depend clause, this is against how everything else has been handled in OpenMP so far (where some identifiers could have special meaning in some OpenMP clauses or pragmas but not elsewhere). The patch handles it by making it a conditional keyword (for -fopenmp only) and emitting a better diagnostics when it is used in a primary expression. Having a nicer diagnostics when e.g. trying to do int omp_all_memory; or int *omp_all_memory[10]; etc. would mean changing too many spots and hooking into name lookups to reject declaring any such symbols would be too ugly and I'm afraid there are way too many spots where one can introduce a name (variables, functions, namespaces, struct, enum, enumerators, template arguments, ...). Otherwise, the handling is quite simple, normal depend clauses lower into addresses of variables being handed over to the library, for omp_all_memory I'm using NULL pointers. omp_all_memory can only be used with inout or out depend kinds and means that a task is dependent on all previously created sibling tasks that have any dependency (of any depend kind) and that any later created sibling tasks will be dependent on it if they have any dependency. 2022-05-12 Jakub Jelinek <jakub@redhat.com> gcc/ * gimplify.cc (gimplify_omp_depend): Don't build_fold_addr_expr if null_pointer_node. (gimplify_scan_omp_clauses): Likewise. * tree-pretty-print.cc (dump_omp_clause): Print null_pointer_node as omp_all_memory. gcc/c-family/ * c-common.h (enum rid): Add RID_OMP_ALL_MEMORY. * c-omp.cc (c_finish_omp_depobj): Don't build_fold_addr_expr if null_pointer_node. gcc/c/ * c-parser.cc (c_parse_init): Register omp_all_memory as keyword if flag_openmp. (c_parser_postfix_expression): Diagnose uses of omp_all_memory in postfix expressions. (c_parser_omp_variable_list): Handle omp_all_memory in depend clause. * c-typeck.cc (c_finish_omp_clauses): Handle omp_all_memory keyword in depend clause as null_pointer_node, diagnose invalid uses. gcc/cp/ * lex.cc (init_reswords): Register omp_all_memory as keyword if flag_openmp. * parser.cc (cp_parser_primary_expression): Diagnose uses of omp_all_memory in postfix expressions. (cp_parser_omp_var_list_no_open): Handle omp_all_memory in depend clause. * semantics.cc (finish_omp_clauses): Handle omp_all_memory keyword in depend clause as null_pointer_node, diagnose invalid uses. * pt.cc (tsubst_omp_clause_decl): Pass through omp_all_memory. gcc/testsuite/ * c-c++-common/gomp/all-memory-1.c: New test. * c-c++-common/gomp/all-memory-2.c: New test. * c-c++-common/gomp/all-memory-3.c: New test. * g++.dg/gomp/all-memory-1.C: New test. * g++.dg/gomp/all-memory-2.C: New test. libgomp/ * libgomp.h (struct gomp_task): Add depend_all_memory member. * task.c (gomp_init_task): Initialize depend_all_memory. (gomp_task_handle_depend): Handle omp_all_memory. (gomp_task_run_post_handle_depend_hash): Clear parent->depend_all_memory if equal to current task. (gomp_task_maybe_wait_for_dependencies): Handle omp_all_memory. * testsuite/libgomp.c-c++-common/depend-1.c: New test. * testsuite/libgomp.c-c++-common/depend-2.c: New test. * testsuite/libgomp.c-c++-common/depend-3.c: New test. (cherry picked from commit 7f78783dbedca0183d193e475262ca3c489fd365)
2022-06-28libgomp: Remove unused '--with-hsa-runtime', '--with-hsa-runtime-include', ↵Thomas Schwinge2-12/+8
'--with-hsa-runtime-lib' With recent commit 2e309a4eff80e55b53d32d26926a2a94eabfea21 "libgomp testsuite: Don't amend 'LD_LIBRARY_PATH' for system-provided HSA Runtime library", and commit d6adba307508c75f1ccb2121eb1a43c9ab1d4056 "libgomp GCN plugin: Clean up unused references to system-provided HSA Runtime library", the last uses of '--with-hsa-runtime' etc. are gone. gcc/ * doc/install.texi: Don't document '--with-hsa-runtime', '--with-hsa-runtime-include', '--with-hsa-runtime-lib'. libgomp/ * plugin/configfrag.ac: Remove '--with-hsa-runtime', '--with-hsa-runtime-include', '--with-hsa-runtime-lib' processing. * Makefile.in: Regenerate. * configure: Likewise. * testsuite/Makefile.in: Likewise. (cherry picked from commit 876ac21b7e796f9efb859dfb46ae2a4126b0b782)
2022-06-28OpenMP, libgomp: Add new runtime routine omp_target_is_accessible.Marcel Vollweiler2-0/+9
gcc/ChangeLog: * omp-low.cc (omp_runtime_api_call): Added target_is_accessible to omp_runtime_apis array. libgomp/ChangeLog: * libgomp.map: Added omp_target_is_accessible. * libgomp.texi: Tagged omp_target_is_accessible as supported. * omp.h.in: Added omp_target_is_accessible. * omp_lib.f90.in: Added interface for omp_target_is_accessible. * omp_lib.h.in: Likewise. * target.c (omp_target_is_accessible): Added implementation of omp_target_is_accessible. * testsuite/libgomp.c-c++-common/target-is-accessible-1.c: New test. * testsuite/libgomp.fortran/target-is-accessible-1.f90: New test. (cherry picked from commit 4043f53cb4a3541f7a6e4f4132419f78ab7ec4f7)
2022-06-28Fortran: Add support for OMP non-rectangular loops.Sandra Loosemore13-25/+709
This patch adds support for OMP 5.1 "canonical loop nest form" to the Fortran front end, marks non-rectangular loops for processing by the middle end, and implements missing checks in the gimplifier for additional prohibitions on non-rectangular loops. Note that the OMP spec also prohibits non-rectangular loops with the TILE construct; that construct hasn't been implemented yet, so that error will need to be filled in later. gcc/fortran/ * gfortran.h (struct gfc_omp_clauses): Add non_rectangular bit. * openmp.cc (is_outer_iteration_variable): New function. (expr_is_invariant): New function. (bound_expr_is_canonical): New function. (resolve_omp_do): Replace existing non-rectangularity error with check for canonical form and setting non_rectangular bit. * trans-openmp.cc (gfc_trans_omp_do): Transfer non_rectangular flag to generated tree structure. gcc/ * gimplify.cc (gimplify_omp_for): Update messages for SCHEDULED and ORDERED clause conflict errors. Add check for GRAINSIZE and NUM_TASKS on TASKLOOP. gcc/testsuite/ * c-c++-common/gomp/loop-6.c (f3): New function to test TASKLOOP diagnostics. * gfortran.dg/gomp/collapse1.f90: Update expected messages. * gfortran.dg/gomp/pr85313.f90: Remove dg-error on non-rectangular loops that are now accepted. * gfortran.dg/gomp/non-rectangular-loop.f90: New file. * gfortran.dg/gomp/canonical-loop-1.f90: New file. * gfortran.dg/gomp/canonical-loop-2.f90: New file. (cherry picked from commit 705bcedf6eae2d7c68bd3df2c98dad4f06650fde)
2022-06-28OpenMP, libgomp: Add new runtime routine omp_get_mapped_ptr.Marcel Vollweiler2-0/+9
This patch adds the OpenMP runtime routine "omp_get_mapped_ptr" which was introduced in OpenMP 5.1. gcc/ChangeLog: * omp-low.cc (omp_runtime_api_call): Added get_mapped_ptr to omp_runtime_apis array. libgomp/ChangeLog: * libgomp.map: Added omp_get_mapped_ptr. * libgomp.texi: Tagged omp_get_mapped_ptr as supported. * omp.h.in: Added omp_get_mapped_ptr. * omp_lib.f90.in: Added interface for omp_get_mapped_ptr. * omp_lib.h.in: Likewise. * target.c (omp_get_mapped_ptr): Added implementation of omp_get_mapped_ptr. * testsuite/libgomp.c-c++-common/get-mapped-ptr-1.c: New test. * testsuite/libgomp.c-c++-common/get-mapped-ptr-2.c: New test. * testsuite/libgomp.c-c++-common/get-mapped-ptr-3.c: New test. * testsuite/libgomp.c-c++-common/get-mapped-ptr-4.c: New test. * testsuite/libgomp.fortran/get-mapped-ptr-1.f90: New test. * testsuite/libgomp.fortran/get-mapped-ptr-2.f90: New test. * testsuite/libgomp.fortran/get-mapped-ptr-3.f90: New test. * testsuite/libgomp.fortran/get-mapped-ptr-4.f90: New test. (cherry picked from commit 941cdc8b6d29f9fe494fdd244e96a5e5aa08ba32)
2022-06-28Fortran: Add location info to OpenMP tree nodesSandra Loosemore2-0/+29
gcc/fortran/ * trans-openmp.cc (gfc_trans_omp_critical): Set location on OMP tree node. (gfc_trans_omp_do): Likewise. (gfc_trans_omp_masked): Likewise. (gfc_trans_omp_do_simd): Likewise. (gfc_trans_omp_scope): Likewise. (gfc_trans_omp_taskgroup): Likewise. (gfc_trans_omp_taskwait): Likewise. (gfc_trans_omp_distribute): Likewise. (gfc_trans_omp_taskloop): Likewise. (gfc_trans_omp_master_masked_taskloop): Likewise. (cherry picked from commit 198bd0d599e0a91df1cfa6ab37545d05dff48e97)
2022-06-28openmp: Implement uses_allocators clauseChung-Lin Tang33-16/+1758
This is a merge of: https://gcc.gnu.org/pipermail/gcc-patches/2022-June/596412.html For user defined allocator handles, this allows target regions to assign memory space and traits to allocators, and automatically calls omp_init/destroy_allocator() in the beginning/end of the target region. For pre-defined allocators (i.e. omp_..._mem_alloc names), this is a no-op, such clauses are not created. Asides from the front-end portions, the target region transforms are done in gimplify_omp_workshare. This patch also includes added changes to enforce the "allocate allocator must be also in a uses_allocator clause". This is done during gimplify_scan_omp_clauses. gcc/c-family/ChangeLog: * c-omp.cc (c_omp_split_clauses): Add OMP_CLAUSE_USES_ALLOCATORS case. * c-pragma.h (enum pragma_omp_clause): Add PRAGMA_OMP_CLAUSE_USES_ALLOCATORS. gcc/c/ChangeLog: * c-parser.cc (c_parser_omp_clause_name): Add case for uses_allocators clause. (c_parser_omp_clause_uses_allocators): New function. (c_parser_omp_all_clauses): Add PRAGMA_OMP_CLAUSE_USES_ALLOCATORS case. (OMP_TARGET_CLAUSE_MASK): Add PRAGMA_OMP_CLAUSE_USES_ALLOCATORS to mask. * c-typeck.cc (c_finish_omp_clauses): Add case handling for OMP_CLAUSE_USES_ALLOCATORS. gcc/cp/ChangeLog: * parser.cc (cp_parser_omp_clause_name): Add case for uses_allocators clause. (cp_parser_omp_clause_uses_allocators): New function. (cp_parser_omp_all_clauses): Add PRAGMA_OMP_CLAUSE_USES_ALLOCATORS case. (OMP_TARGET_CLAUSE_MASK): Add PRAGMA_OMP_CLAUSE_USES_ALLOCATORS to mask. * semantics.cc (finish_omp_clauses): Add case handling for OMP_CLAUSE_USES_ALLOCATORS. fortran/ChangeLog: * gfortran.h (struct gfc_omp_namelist): Add memspace_sym, traits_sym fields. (OMP_LIST_USES_ALLOCATORS): New list enum. * openmp.cc (enum omp_mask2): Add OMP_CLAUSE_USES_ALLOCATORS. (gfc_match_omp_clause_uses_allocators): New function. (gfc_match_omp_clauses): Add case to handle OMP_CLAUSE_USES_ALLOCATORS. (OMP_TARGET_CLAUSES): Add OMP_CLAUSE_USES_ALLOCATORS. (resolve_omp_clauses): Add "USES_ALLOCATORS" to clause_names[]. * dump-parse-tree.cc (show_omp_namelist): Handle OMP_LIST_USES_ALLOCATORS. (show_omp_clauses): Likewise. * trans-array.cc (gfc_conv_array_initializer): Adjust array index to always be a created tree expression instead of NULL_TREE when zero. * trans-openmp.cc (gfc_trans_omp_clauses): For ALLOCATE clause, handle using gfc_trans_omp_variable for EXPR_VARIABLE exprs. Add handling of OMP_LIST_USES_ALLOCATORS case. * types.def (BT_FN_VOID_PTRMODE): Define. (BT_FN_PTRMODE_PTRMODE_INT_PTR): Define. gcc/ChangeLog: * builtin-types.def (BT_FN_VOID_PTRMODE): Define. (BT_FN_PTRMODE_PTRMODE_INT_PTR): Define. * omp-builtins.def (BUILT_IN_OMP_INIT_ALLOCATOR): Define. (BUILT_IN_OMP_DESTROY_ALLOCATOR): Define. * tree-core.h (enum omp_clause_code): Add OMP_CLAUSE_USES_ALLOCATORS. * tree-pretty-print.cc (dump_omp_clause): Handle OMP_CLAUSE_USES_ALLOCATORS. * tree.h (OMP_CLAUSE_USES_ALLOCATORS_ALLOCATOR): New macro. (OMP_CLAUSE_USES_ALLOCATORS_MEMSPACE): New macro. (OMP_CLAUSE_USES_ALLOCATORS_TRAITS): New macro. * tree.cc (omp_clause_num_ops): Add OMP_CLAUSE_USES_ALLOCATORS. (omp_clause_code_name): Add "uses_allocators". (walk_tree_1): Add OMP_CLAUSE_USES_ALLOCATORS case. * gimplify.cc (gimplify_scan_omp_clauses): Add checking of OpenMP target region allocate clauses, to require a uses_allocators clause to exist for allocators. (gimplify_omp_workshare): Add handling of OMP_CLAUSE_USES_ALLOCATORS for OpenMP target regions; create calls of omp_init/destroy_allocator around target region body. * omp-low.cc (lower_private_allocate): Adjust receiving of allocator. (lower_rec_input_clauses): Likewise. (create_task_copyfn): Add dereference for allocator if needed. * system.h (startswith): New function. gcc/testsuite/ChangeLog: * c-c++-common/gomp/uses_allocators-1.c: New test. * c-c++-common/gomp/uses_allocators-2.c: New test. * c-c++-common/gomp/uses_allocators-3.c: New test. * gfortran.dg/gomp/allocate-1.f90: Adjust testcase. * gfortran.dg/gomp/uses_allocators-1.f90: New test. * gfortran.dg/gomp/uses_allocators-2.f90: New test. * gfortran.dg/gomp/uses_allocators-3.f90: New test.
2022-06-28amdgcn: Add gfx90a supportAndrew Stubbs10-23/+128
This adds architecture options and multilibs for the AMD GFX90a GPUs. It also tidies up some of the ISA selection code, and corrects a few small mistake in the gfx908 naming. gcc/ChangeLog: * config.gcc (amdgcn): Accept --with-arch=gfx908 and gfx90a. * config/gcn/gcn-opts.h (enum gcn_isa): New. (TARGET_GCN3): Use enum gcn_isa. (TARGET_GCN3_PLUS): Likewise. (TARGET_GCN5): Likewise. (TARGET_GCN5_PLUS): Likewise. (TARGET_CDNA1): New. (TARGET_CDNA1_PLUS): New. (TARGET_CDNA2): New. (TARGET_CDNA2_PLUS): New. (TARGET_M0_LDS_LIMIT): New. (TARGET_PACKED_WORK_ITEMS): New. * config/gcn/gcn.cc (gcn_isa): Change to enum gcn_isa. (gcn_option_override): Recognise CDNA ISA variants. (gcn_omp_device_kind_arch_isa): Support gfx90a. (gcn_expand_prologue): Make m0 init optional. Add support for packed work items. (output_file_start): Support gfx90a. (gcn_hsa_declare_function_name): Support gfx90a metadata. * config/gcn/gcn.h (TARGET_CPU_CPP_BUILTINS):Add __CDNA1__ and __CDNA2__. * config/gcn/gcn.md (<su>mulsi3_highpart): Use TARGET_GCN5_PLUS. (<su>mulsi3_highpart_imm): Likewise. (<su>mulsidi3): Likewise. (<su>mulsidi3_imm): Likewise. * config/gcn/gcn.opt (gpu_type): Add gfx90a. * config/gcn/mkoffload.cc (EF_AMDGPU_MACH_AMDGCN_GFX90a): New. (main): Support gfx90a. * config/gcn/t-gcn-hsa: Add gfx90a multilib. * config/gcn/t-omp-device: Add gfx90a isa. libgomp/ChangeLog: * plugin/plugin-gcn.c (EF_AMDGPU_MACH): Add EF_AMDGPU_MACH_AMDGCN_GFX90a. (gcn_gfx90a_s): New. (isa_hsa_name): Support gfx90a. (isa_code): Likewise. (cherry picked from commit cde52d3a2d02d037da53e6974d5e39021030b346)
2022-06-28amdgcn: Remove LLVM 9 assembler/linker supportAndrew Stubbs8-812/+105
The minimum required LLVM version is now 13.0.1, and is enforced by configure. gcc/ChangeLog: * config.in: Regenerate. * config/gcn/gcn-hsa.h (X_FIJI): Delete. (X_900): Delete. (X_906): Delete. (X_908): Delete. (S_FIJI): Delete. (S_900): Delete. (S_906): Delete. (S_908): Delete. (NO_XNACK): New macro. (NO_SRAM_ECC): New macro. (SRAMOPT): Keep only v4 variant. (HSACO3_SELECT_OPT): Delete. (DRIVER_SELF_SPECS): Delete. (ASM_SPEC): Remove LLVM 9 support. * config/gcn/gcn-valu.md (gather<mode>_insn_2offsets<exec>): Remove assembler bug workaround. (scatter<mode>_insn_2offsets<exec_scatter>): Likewise. * config/gcn/gcn.cc (output_file_start): Remove LLVM 9 support. (print_operand_address): Remove assembler bug workaround. * config/gcn/mkoffload.cc (EF_AMDGPU_XNACK_V3): Delete. (EF_AMDGPU_SRAM_ECC_V3): Delete. (SET_XNACK_ON): Delete v3 variants. (SET_XNACK_OFF): Delete v3 variants. (TEST_XNACK): Delete v3 variants. (SET_SRAM_ECC_ON): Delete v3 variants. (SET_SRAM_ECC_ANY): Delete v3 variants. (SET_SRAM_ECC_OFF): Delete v3 variants. (SET_SRAM_ECC_UNSUPPORTED): Delete v3 variants. (TEST_SRAM_ECC_ANY): Delete v3 variants. (TEST_SRAM_ECC_ON): Delete v3 variants. (copy_early_debug_info): Remove v3 support. (main): Remove v3 support. * configure: Regenerate. * configure.ac: Replace all GCN feature checks with a version check. (cherry picked from commit 8086230e7ac619c0b0eeb6e15df7975ac214725f)
2022-06-28OpenMP: Handle descriptors in target's firstprivate [PR104949]Tobias Burnus9-11/+188
For allocatable/pointer arrays, a firstprivate to a device not only needs to privatize the descriptor but also the actual data. This is implemented as: firstprivate(x) firstprivate(x.data) attach(x [bias: &x.data-&x) where the address of x in device memory is saved in hostaddrs[i] by libgomp and the middle end actually passes hostaddrs[i]' to attach. As side effect, has_device_addr(array_desc) had to be changed: before, it was converted to firstprivate in the front end; now it is handled in omp-low.cc as has_device_addr requires a shallow firstprivate (not touching the data pointer) while the normal firstprivate requires (now) a deep firstprivate. gcc/fortran/ChangeLog: PR fortran/104949 * f95-lang.cc (LANG_HOOKS_OMP_ARRAY_SIZE): Redefine. * trans-openmp.cc (gfc_omp_array_size): New. (gfc_trans_omp_variable_list): Never turn has_device_addr to firstprivate. * trans.h (gfc_omp_array_size): New. gcc/ChangeLog: PR fortran/104949 * langhooks-def.h (lhd_omp_array_size): New. (LANG_HOOKS_OMP_ARRAY_SIZE): Define. (LANG_HOOKS_DECLS): Add it. * langhooks.cc (lhd_omp_array_size): New. * langhooks.h (struct lang_hooks_for_decls): Add hook. * omp-low.cc (scan_sharing_clauses, lower_omp_target): Handle GOMP_MAP_FIRSTPRIVATE for array descriptors. libgomp/ChangeLog: PR fortran/104949 * target.c (gomp_map_vars_internal, copy_firstprivate_data): Support attach for GOMP_MAP_FIRSTPRIVATE. * testsuite/libgomp.fortran/target-firstprivate-1.f90: New test. * testsuite/libgomp.fortran/target-firstprivate-2.f90: New test. * testsuite/libgomp.fortran/target-firstprivate-3.f90: New test. (cherry picked from commit 49d1a2f91325fa8cc011149e27e5093a988b3a49)
2022-06-28Fortran: Fix proc pointer as elemental arg handlingTobias Burnus4-0/+16
The vtab's _callback function calls the elemental 'cb' cb (var(:)%comp, comp_types_vtable._callback); which gets called in a scalarization loop as 'var' might be a nonscalar. Without the patch, that got translated as: D.1234 = &comp_types_vtable._callback ... cb (&(*D.4060)[S.3 + D.4071], &D.1234); where 'D.1234' is function_type. With the patch, it remains a pointer; i.e. D.1234 = comp... and 'cb (..., D.1234)', avoiding ME ICE. Note: Fortran (F2018, C15100) requires that dummy arguments are dummy data objects, which rules out dummy procs/proc-pointer dummies, which is enforced in resolve_fl_procedure. Thus, this change only affects the internally generated code. gcc/fortran/ChangeLog: * trans-array.cc (gfc_scalar_elemental_arg_saved_as_reference): Return true for attr.proc_pointer expressions. gcc/testsuite/ChangeLog: * gfortran.dg/finalize_38.f90: Compile with -Ofast.
2022-06-28Fortran: Fix finalization resolution with deep copy (cont)Tobias Burnus6-13/+192
gcc/fortran/ChangeLog: * resolve.cc (gfc_resolve_finalizers): Remove gfc_resolve_finalizers calls, use gfc_is_finalizable. (resolve_fl_derived): Resolve derived-type components first. gcc/testsuite/ChangeLog: * gfortran.dg/abstract_type_6.f03: Remove dg-error as now hidden by other errors; copy to ... * gfortran.dg/abstract_type_6a.f03: ... here; remove some error to diagnose the error. * gfortran.dg/finalize_39.f90: New test.
2022-06-28Fortran: Fix finalization resolution with deep copyTobias Burnus4-3/+151
Follow-up patch to "Fortran/OpenMP: Support mapping of DT with allocatable components" https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591144.html gcc/fortran/ChangeLog: * resolve.cc (gfc_resolve_finalizers): Also resolve allocatable comps. gcc/testsuite/ChangeLog: * gfortran.dg/finalize_38.f90: New test.
2022-06-28OpenMP: Fix use_device_{addr,ptr} with in-data-sharing argTobias Burnus2-9/+17
For array-descriptor vars, the descriptor is assigned to a temporary. However, this failed when the clause's argument was in turn in a data-sharing clause as the outer context's VALUE_EXPR wasn't used. gcc/ChangeLog: * omp-low.cc (lower_omp_target): Fix use_device_{addr,ptr} with list item that is in an outer data-sharing clause. libgomp/ChangeLog: * testsuite/libgomp.fortran/use_device_addr-5.f90: New test. (cherry picked from commit 3f8c389fe90bf565a6221a46bb7fb745dd4c1510)
2022-06-28openmp: unified_address supportAndrew Stubbs11-23/+82
This makes "requires unified_address" work by making it eqivalent to "requires unified_shared_memory". This is more than is strictly necessary, but should be standard compliant. gcc/c/ChangeLog: * c-parser.cc (c_parser_omp_requires): Check requires unified_address for conflict with -foffload-memory=shared. gcc/cp/ChangeLog: * parser.cc (cp_parser_omp_requires): Check requires unified_address for conflict with -foffload-memory=shared. gcc/fortran/ChangeLog: * openmp.cc (gfc_match_omp_requires): Check requires unified_address for conflict with -foffload-memory=shared. gcc/ChangeLog: * omp-low.cc: Do USM transformations for "unified_address". gcc/testsuite/ChangeLog: * c-c++-common/gomp/usm-4.c: New test. * gfortran.dg/gomp/usm-4.f90: New test.
2022-06-28openmp: Do USM transform for omp_target_allocAndrew Stubbs5-4/+30
OpenMP 5.0 says that omp_target_alloc should return USM addresses. gcc/ChangeLog: * omp-low.cc (usm_transform): Transform omp_target_alloc and omp_target_free. libgomp/ChangeLog: * testsuite/libgomp.c/usm-6.c: Add omp_target_alloc. gcc/testsuite/ChangeLog: * c-c++-common/gomp/usm-2.c: Add omp_target_alloc. * c-c++-common/gomp/usm-3.c: Add omp_target_alloc.
2022-06-28Fix a crash due to mismatch of free and GOMP_alloc.Hafiz Abid Qadeer2-1/+10
With allocate directive, we replace the malloc calls to GOMP_alloc if it is associated with the allocate statement. The memory was supposed to be free-d by the implicitely generated free calls which also get replaced. But if user explicitely deallocated the memory using the deallocate statement, it can cause a mismatch. This commit handles that case and also replaces the free call generated for deallocate clause. Also added deallocate in the testcase and tidied it up a bit. gcc/ChangeLog: * omp-low.cc (lower_omp_allocate): Move allocate declaration inside loop. Set it to false at the end of condition. libgomp/ChangeLog: * testsuite/libgomp.fortran/allocate-2.f90: Remove commented lines. Add deallocate. Remove omp_atk_pool_size trait.
2022-06-28openmp: BUILT_IN_GOMP_ENABLE_PINNED_MODEAndrew Stubbs3-3/+10
Rework the GOMP_enable_pinned_mode call so that it works on powerpc where the old way gave a local call. gcc/ChangeLog: * omp-builtins.def (BUILT_IN_GOMP_ENABLE_PINNED_MODE): New. * omp-low.cc (omp_enable_pinned_mode): Use BUILT_IN_GOMP_ENABLE_PINNED_MODE.
2022-06-28openmp: -foffload-memory=pinnedAndrew Stubbs4-0/+110
Implement the -foffload-memory=pinned option such that libgomp is instructed to enable fully-pinned memory at start-up. The option is intended to provide a performance boost to certain offload programs without modifying the code. This feature only works on Linux, at present, and simply calls mlockall to enable always-on memory pinning. It requires that the ulimit feature is set high enough to accommodate all the program's memory usage. In this mode the ompx_pinned_memory_alloc feature is disabled as it is not needed and may conflict. Backport of the patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591354.html gcc/ChangeLog: * omp-low.cc (omp_enable_pinned_mode): New function. (execute_lower_omp): Call omp_enable_pinned_mode. libgomp/ChangeLog: * config/linux/allocator.c (always_pinned_mode): New variable. (GOMP_enable_pinned_mode): New function. (linux_memspace_alloc): Disable pinning when always_pinned_mode set. (linux_memspace_calloc): Likewise. (linux_memspace_free): Likewise. (linux_memspace_realloc): Likewise. * libgomp.map (GOMP_5.1.1): New version space with GOMP_enable_pinned_mode. * testsuite/libgomp.c/alloc-pinned-7.c: New test. gcc/testsuite/ChangeLog: * c-c++-common/gomp/alloc-pinned-1.c: New test.
2022-06-28openmp: Use libgomp memory allocation functions with unified shared memory.Hafiz Abid Qadeer12-0/+402
This patches changes calls to malloc/free/calloc/realloc/aligned_alloc and operator new to memory allocation functions in libgomp with allocator=ompx_unified_shared_mem_alloc. This helps existing code to benefit from the unified shared memory. The libgomp does the correct thing with all the mapping constructs and there is no memory copies if the pointer is pointing to unified shared memory. We only replace replacable new operator and not the class member or placement new. Backport of a patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591353.html gcc/ChangeLog: * omp-low.cc (usm_transform): New function. (make_pass_usm_transform): Likewise. (class pass_usm_transform): New. * passes.def: Add pass_usm_transform. * tree-pass.h (make_pass_usm_transform): New declaration. gcc/testsuite/ChangeLog: * c-c++-common/gomp/usm-2.c: New test. * c-c++-common/gomp/usm-3.c: New test. * g++.dg/gomp/usm-1.C: New test. * g++.dg/gomp/usm-2.C: New test. * g++.dg/gomp/usm-3.C: New test. * gfortran.dg/gomp/usm-2.f90: New test. * gfortran.dg/gomp/usm-3.f90: New test. libgomp/ChangeLog: * testsuite/libgomp.c/usm-6.c: New test. * testsuite/libgomp.c++/usm-1.C: Likewise.
2022-06-28openmp: allow requires unified_shared_memoryAndrew Stubbs9-0/+66
This is the front-end portion of the Unified Shared Memory implementation. It checks that -foffload-memory isn't set to an incompatible mode. Backport of the patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591351.html gcc/c/ChangeLog: * c-parser.cc (c_parser_omp_requires): Check compatibility of -foffload-memory option with requires directive. gcc/cp/ChangeLog: * parser.cc (cp_parser_omp_requires): Check compatibility of -foffload-memory option with requires directive. gcc/fortran/ChangeLog: * openmp.cc (gfc_match_omp_requires): Check compatibility of -foffload-memory option with requires directive. gcc/testsuite/ChangeLog: * c-c++-common/gomp/usm-1.c: New test. * gfortran.dg/gomp/usm-1.f90: New test.
2022-06-28openmp: Add -foffload-memoryAndrew Stubbs4-1/+47
Add a new option. It will be used in follow-up patches. Backport of the patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-March/591350.html gcc/ChangeLog: * common.opt: Add -foffload-memory and its enum values. * coretypes.h (enum offload_memory): New. * doc/invoke.texi: Document -foffload-memory.
2022-06-28Lower allocate directive (OpenMP 5.0).Hafiz Abid Qadeer6-0/+192
Backport of a patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-January/588372.html This patch looks for malloc/free calls that were generated by allocate statement that is associated with allocate directive and replaces them with GOMP_alloc and GOMP_free. gcc/ChangeLog: * omp-low.cc (scan_sharing_clauses): Handle OMP_CLAUSE_ALLOCATOR. (scan_omp_allocate): New. (scan_omp_1_stmt): Call it. (lower_omp_allocate): New function. (lower_omp_1): Call it. gcc/testsuite/ChangeLog: * gfortran.dg/gomp/allocate-6.f90: Add tests. * gfortran.dg/gomp/allocate-7.f90: New test. * gfortran.dg/gomp/allocate-8.f90: New test. libgomp/ChangeLog: * testsuite/libgomp.fortran/allocate-2.f90: New test.
2022-06-28Gimplify allocate directive (OpenMP 5.0).Hafiz Abid Qadeer10-4/+205
Backport of a patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-January/588371.html gcc/ChangeLog: * doc/gimple.texi: Describe GIMPLE_OMP_ALLOCATE. * gimple-pretty-print.cc (dump_gimple_omp_allocate): New function. (pp_gimple_stmt_1): Call it. * gimple.cc (gimple_build_omp_allocate): New function. * gimple.def (GIMPLE_OMP_ALLOCATE): New node. * gimple.h (enum gf_mask): Add GF_OMP_ALLOCATE_KIND_MASK, GF_OMP_ALLOCATE_KIND_ALLOCATE and GF_OMP_ALLOCATE_KIND_FREE. (struct gomp_allocate): New. (is_a_helper <gomp_allocate *>::test): New. (is_a_helper <const gomp_allocate *>::test): New. (gimple_build_omp_allocate): Declare. (gimple_omp_subcode): Replace GIMPLE_OMP_TEAMS with GIMPLE_OMP_ALLOCATE. (gimple_omp_allocate_set_clauses): New. (gimple_omp_allocate_set_kind): Likewise. (gimple_omp_allocate_clauses): Likewise. (gimple_omp_allocate_kind): Likewise. (CASE_GIMPLE_OMP): Add GIMPLE_OMP_ALLOCATE. * gimplify.cc (gimplify_omp_allocate): New. (gimplify_expr): Call it. * gsstruct.def (GSS_OMP_ALLOCATE): Define. gcc/testsuite/ChangeLog: * gfortran.dg/gomp/allocate-6.f90: Add tests.
2022-06-28Handle cleanup of omp allocated variables (OpenMP 5.0).Hafiz Abid Qadeer11-1/+106
Currently we are only handling omp allocate directive that is associated with an allocate statement. This statement results in malloc and free calls. The malloc calls are easy to get to as they are in the same block as allocate directive. But the free calls come in a separate cleanup block. To help any later passes finding them, an allocate directive is generated in the cleanup block with kind=free. The normal allocate directive is given kind=allocate. Backport of a patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-January/588370.html gcc/fortran/ChangeLog: * gfortran.h (struct access_ref): Declare new members omp_allocated and omp_allocated_end. * openmp.cc (gfc_match_omp_allocate): Set new_st.resolved_sym to NULL. (prepare_omp_allocated_var_list_for_cleanup): New function. (gfc_resolve_omp_allocate): Call it. * trans-decl.cc (gfc_trans_deferred_vars): Process omp_allocated. * trans-openmp.cc (gfc_trans_omp_allocate): Set kind for the stmt generated for allocate directive. gcc/ChangeLog: * tree-core.h (struct tree_base): Add comments. * tree-pretty-print.cc (dump_generic_node): Handle allocate directive kind. * tree.h (OMP_ALLOCATE_KIND_ALLOCATE): New define. (OMP_ALLOCATE_KIND_FREE): Likewise. gcc/testsuite/ChangeLog: * gfortran.dg/gomp/allocate-6.f90: Test kind of allocate directive.
2022-06-28Translate allocate directive (OpenMP 5.0).Hafiz Abid Qadeer10-0/+183
Backport of a patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-January/588369.html gcc/fortran/ChangeLog: * trans-openmp.cc (gfc_trans_omp_clauses): Handle OMP_LIST_ALLOCATOR. (gfc_trans_omp_allocate): New function. (gfc_trans_omp_directive): Handle EXEC_OMP_ALLOCATE. gcc/ChangeLog: * tree-pretty-print.cc (dump_omp_clause): Handle OMP_CLAUSE_ALLOCATOR. (dump_generic_node): Handle OMP_ALLOCATE. * tree.def (OMP_ALLOCATE): New. * tree.h (OMP_ALLOCATE_CLAUSES): Likewise. (OMP_ALLOCATE_DECL): Likewise. (OMP_ALLOCATE_ALLOCATOR): Likewise. * tree.cc (omp_clause_num_ops): Add entry for OMP_CLAUSE_ALLOCATOR. gcc/testsuite/ChangeLog: * gfortran.dg/gomp/allocate-6.f90: New test.
2022-06-28Add parsing support for allocate directive (OpenMP 5.0)Hafiz Abid Qadeer12-5/+435
Currently we only make use of this directive when it is associated with an allocate statement. Backport of a patch posted at https://gcc.gnu.org/pipermail/gcc-patches/2022-January/588368.html gcc/fortran/ChangeLog: * dump-parse-tree.cc (show_omp_node): Handle EXEC_OMP_ALLOCATE. (show_code_node): Likewise. * gfortran.h (enum gfc_statement): Add ST_OMP_ALLOCATE. (OMP_LIST_ALLOCATOR): New enum value. (enum gfc_exec_op): Add EXEC_OMP_ALLOCATE. * match.h (gfc_match_omp_allocate): New function. * openmp.cc (enum omp_mask1): Add OMP_CLAUSE_ALLOCATOR. (OMP_ALLOCATE_CLAUSES): New define. (gfc_match_omp_allocate): New function. (resolve_omp_clauses): Add ALLOCATOR in clause_names. (omp_code_to_statement): Handle EXEC_OMP_ALLOCATE. (EMPTY_VAR_LIST): New define. (check_allocate_directive_restrictions): New function. (gfc_resolve_omp_allocate): Likewise. (gfc_resolve_omp_directive): Handle EXEC_OMP_ALLOCATE. * parse.cc (decode_omp_directive): Handle ST_OMP_ALLOCATE. (next_statement): Likewise. (gfc_ascii_statement): Likewise. * resolve.cc (gfc_resolve_code): Handle EXEC_OMP_ALLOCATE. * st.cc (gfc_free_statement): Likewise. * trans.cc (trans_code): Likewise
2022-06-28Set omp_requires_mask for dynamic_allocators.Hafiz Abid Qadeer2-0/+8
This is backport of a patch posted in https://gcc.gnu.org/pipermail/gcc-patches/2022-February/590655.html This patch fixes an issue that although gfortran accepts 'requires dynamic_allocators', it does not set the omp_requires_mask accordingly. gcc/fortran/ChangeLog: * parse.cc (gfc_parse_file): Set OMP_REQUIRES_DYNAMIC_ALLOCATORS bit in omp_requires_mask.
2022-06-28Add a restriction on allocate clause (OpenMP 5.0)Hafiz Abid Qadeer6-0/+69
This is backport of a patch posted in https://gcc.gnu.org/pipermail/gcc-patches/2022-February/590597.html An allocate clause in target region must specify an allocator unless the compilation unit has requires construct with dynamic_allocators clause. Current implementation of the allocate clause did not check for this restriction. This patch fills that gap. gcc/ChangeLog: * omp-low.cc (omp_maybe_offloaded_ctx): New prototype. (scan_sharing_clauses): Check a restriction on allocate clause. gcc/testsuite/ChangeLog: * c-c++-common/gomp/allocate-2.c: Add tests. * c-c++-common/gomp/allocate-8.c: New test. * gfortran.dg/gomp/allocate-3.f90: Add tests. * gcc.dg/gomp/pr104517.c: Update.
2022-06-28Fortran/OpenMP: Support mapping of DT with allocatable componentsTobias Burnus24-136/+2527
gcc/fortran/ChangeLog: * class.cc (finalization_scalarizer): Mark syms as artificial. (generate_callback_wrapper): New. (gfc_find_derived_vtab): Call it, add _callback comp. * f95-lang.cc (LANG_HOOKS_OMP_DEEP_MAPPING, LANG_HOOKS_OMP_DEEP_MAPPING_P, LANG_HOOKS_OMP_DEEP_MAPPING_CNT): Redeinfe * gfortran.h (gfc_import_iso_c_binding_module, GFC_CLASS_CALLBACK_DEFAULT_FLAG, GFC_CLASS_CALLBACK_VTABLE_FLAG, GFC_CLASS_CB_ALLOCATABLE, GFC_CLASS_CB_POINTER, GFC_CLASS_CB_PROC_POINTER, GFC_CLASS_CB_VTABLE, GFC_CLASS_CB_VPTR): New. * match.cc (select_type_set_tmp): Propagate allocatable property. * module.cc (MOD_VERSION): Bump due to vtab change. (import_iso_c_binding_module): New import_all arg. (gfc_import_iso_c_binding_module): New. (gfc_use_module): Update call. * openmp.cc (resolve_omp_clauses): Accept DT with alloc comps. * resolve.cc (gfc_resolve_formal_arglist, gfc_resolve_intrinsic, resolve_fl_procedure, resolve_types): Permit some violations for internal code. * trans-array.cc (gfc_conv_descriptor_stride_get, gfc_tree_array_size, gfc_full_array_size): Update for GFC_TYPE_ARRAY_AKIND change. (gfc_conv_expr_descriptor): Likewise; permit calling with tree code. * trans-expr.cc (VTABLE_CALLBACK_FIELD): Add. (VTAB_GET_FIELD_GEN): Use it. (VTABLE_DEALLOCATE_FIELD): Undef at the end. (gfc_conv_expr_reference): Fixes; avoid unneccessary temp var. * trans-intrinsic.cc (gfc_conv_intrinsic_sizeof, gfc_conv_associated): Fix class and comp-ref handling. (conv_isocbinding_function): Remove buggy code. * trans-openmp.cc (gfc_has_alloc_comps): Add ptr_ok arg. (gfc_omp_private_outer_ref, gfc_walk_alloc_comps, gfc_omp_clause_default_ctor, gfc_omp_clause_copy_ctor, gfc_omp_clause_assign_op, gfc_omp_clause_dtor, (gfc_omp_finish_clause): Update call. (GFC_MAP_TOKEN_DATA, GFC_MAP_TOKEN_SIZES, GFC_MAP_TOKEN_KINDS, GFC_MAP_TOKEN_DATA_OFFSET, GFC_MAP_TOKEN_OFFSET, GFC_MAP_TOKEN_FLAGS, GFC_MAP_TOKEN_DETACH): Define. (gfc_omp_get_token_data, gfc_omp_get_token_sizes, gfc_omp_get_token_kinds, gfc_omp_get_token_offset_data, gfc_omp_get_token_offset, gfc_omp_get_token_flags, gfc_omp_get_token_detach, gfc_omp_get_map_token_type, gfc_omp_get_cb_type, gfc_omp_gen_deep_map_fn, gfc_omp_deep_mapping_map, gfc_omp_deep_mapping_item, gfc_omp_deep_mapping_comps, gfc_omp_gen_simple_loop, gfc_omp_get_array_size, gfc_omp_elmental_loop, gfc_omp_deep_map_kind_p, gfc_omp_deep_mapping_int_p, gfc_omp_deep_mapping_p, gfc_omp_deep_mapping_do), gfc_omp_deep_mapping_cnt, gfc_omp_deep_mapping): New. (gfc_trans_omp_array_section): Save clause decl to survive gimplifying. (gfc_trans_omp_clauses): Likewise; fixes. * trans-types.cc (gfc_build_array_type, gfc_get_derived_type, gfc_get_array_descr_info): Update array kind to distinguish different assumed-rank arrays. * trans.h (gfc_class_vtab_callback_get, gfc_omp_deep_mapping_p, gfc_omp_deep_mapping_cnt, gfc_omp_deep_mapping): New prototypes. (enum gfc_array_kind): Additional GFC_ARRAY_ASSUMED_RANK_* entries. gcc/ChangeLog: * langhooks-def.h (lhd_omp_deep_mapping_p, lhd_omp_deep_mapping_cnt, lhd_omp_deep_mapping): New. (LANG_HOOKS_OMP_DEEP_MAPPING_P, LANG_HOOKS_OMP_DEEP_MAPPING_CNT, LANG_HOOKS_OMP_DEEP_MAPPING): Define. (LANG_HOOKS_DECLS): Use it. * langhooks.cc (lhd_omp_deep_mapping_p, lhd_omp_deep_mapping_cnt, lhd_omp_deep_mapping): New stubs. * langhooks.h (struct lang_hooks_for_decls): Add new hooks * omp-expand.cc (expand_omp_target): Handle dynamic-size addr/sizes/kinds arrays. * omp-low.cc (build_sender_ref, fixup_child_record_type, scan_sharing_clauses, lower_omp_target): Update to handle new hooks and dynamic-size addr/sizes/kinds arrays. libgomp/ChangeLog: * testsuite/libgomp.fortran/allocatable-comp.f90: New test. * testsuite/libgomp.fortran/map-alloc-comp-3.f90: New test. * testsuite/libgomp.fortran/map-alloc-comp-4.f90: New test. * testsuite/libgomp.fortran/map-alloc-comp-5.f90: New test. * testsuite/libgomp.fortran/map-alloc-comp-6.f90: New test. * testsuite/libgomp.fortran/map-alloc-comp-7.f90: New test. gcc/testsuite/ChangeLog: * gfortran.dg/c_loc_test_22.f90: Update scan-tree. * gfortran.dg/finalize_21.f90: Likewise. * gfortran.dg/gomp/map-alloc-comp-1.f90: Remove sorry dg-error.
2022-06-28openmp: Handle C/C++ array reference base-pointers in array sectionsChung-Lin Tang8-4/+79
In cases where a program constructs its own deep-copying for arrays-of-pointers, e.g: #pragma omp target enter data map(to:level->vectors[:N]) for (i = 0; i < N; i++) #pragma omp target enter data map(to:level->vectors[i][:N]) We need to treat the part of the array reference before the array section as a base-pointer (here 'level->vectors[i]'), providing pointer-attachment behavior. This patch adds this inside handle_omp_array_sections(), tracing the whole sequence of array dimensions, creating a whole base-pointer reference iteratively using build_array_ref(). The conditions are that each of the "absorbed" dimensions must be length==1, and the final reference must be of pointer-type (so that pointer attachment makes sense). Merged from: https://gcc.gnu.org/pipermail/gcc-patches/2022-February/590658.html 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/ChangeLog: * gimplify.cc (gimplify_scan_omp_clauses): Add case for attach/detach map kind for ARRAY_REF of POINTER_TYPE. 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.
2022-06-28openmp: Improve handling of nested OpenMP metadirectives in C and C++Kwok Cheung Yeung6-42/+78
This patch fixes a misparsing issue when encountering code like: #pragma omp metadirective when {<selector_set>={...}: A) #pragma omp metadirective when (<selector_set>={...}: B) When called for the first metadirective, analyze_metadirective_body would stop just before the colon in the second metadirective because it naively assumes that the '}' marks the end of a code block. The assertion for clauses to end parsing at the same point is now disabled if a parse error has occurred during the parsing of the clause, since some tokens may not be consumed if a parse error cuts parsing short. 2022-02-18 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/c/ * c-parser.cc (c_parser_omp_construct): Move handling of PRAGMA_OMP_METADIRECTIVE from here... (c_parser_pragma): ...to here. (analyze_metadirective_body): Check that the bracket nesting level is also zero before stopping the adding of tokens on encountering a close brace. (c_parser_omp_metadirective): Modify function signature and update. Do not assert on remaining tokens if there has been a parse error. gcc/cp/ * parser.cc (cp_parser_omp_construct): Move handling of PRAGMA_OMP_METADIRECTIVE from here... (cp_parser_pragma): ...to here. (analyze_metadirective_body): Check that the bracket nesting level is also zero before stopping the adding of tokens on encountering a close brace. (cp_parser_omp_metadirective): Modify function signature and update. Do not assert on remaining tokens if there has been a parse error. gcc/testsuite/ * c-c++-common/gomp/metadirective-1.c (f): Add test for improperly nested metadirectives.
2022-06-28openmp: More Fortran front-end fixes for metadirectivesKwok Cheung Yeung6-2/+37
This adds a check for declarative OpenMP directives in metadirective variants (already present in the C/C++ front-ends), and fixes an ICE when an empty metadirective (i.e. just '!$omp metadirective') is presented. 2022-02-11 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/fortran/ * gfortran.h (is_omp_declarative_stmt): New. * openmp.cc (match_omp_metadirective): Reject declarative OpenMP directives with 'sorry'. * parse.cc (parse_omp_metadirective_body): Check that state stack head is non-null before dereferencing. (is_omp_declarative_stmt): New. gcc/testsuite/ * gfortran.dg/gomp/metadirective-2.f90 (main): Test empty metadirective.
2022-06-28openmp: Eliminate non-matching metadirective variants early in Fortran front-endKwok Cheung Yeung9-16/+101
This patch checks during parsing if a metadirective selector is both resolvable and non-matching - if so, it is removed from further consideration. This is both more efficient, and avoids spurious syntax errors caused by considering combinations of selectors that lead to invalid combinations of OpenMP directives, when that combination would never arise in the first place. This exposes another bug - when metadirectives that are not of the begin-end variety are nested, we might have to drill up through multiple layers of the state stack to reach the state for the next statement. This is now fixed. 2022-02-11 Kwok Cheung Yeung <kcy@codesourcery.com> gcc/ * omp-general.cc (DELAY_METADIRECTIVES_AFTER_LTO): Check that cfun is non-null before derefencing. gcc/fortran/ * decl.cc (gfc_match_end): Search for first previous state that is not COMP_OMP_METADIRECTIVE. * gfortran.h (gfc_skip_omp_metadirective_clause): Add prototype. * openmp.cc (match_omp_metadirective): Skip clause if result of gfc_skip_omp_metadirective_clause is true. * trans-openmp.cc (gfc_trans_omp_set_selector): Add argument and disable expression conversion if false. (gfc_skip_omp_metadirective_clause): New. gcc/testsuite/ * gfortran.dg/gomp/metadirective-8.f90: New.
generated by cgit v1.1 at 2025-09-23 11:19:34 +0000