| Age | Commit message (Collapse) | Author | Files | Lines |
|---|
|
This patch supports "lvalue" parsing (or "locator list item type" parsing)
for several OpenMP clause types for C++, as required for OpenMP 5.0
and above.
This version has been rebased -- some things have changed around
template handling recently, e.g. removal of build_non_dependent_expr and
tsubst_copy. A new potential corner-case issue has shown up regarding
implicit mapping of references to pointer to pointers -- an interaction
with the post-review fixes/rework for the patch here:
https://gcc.gnu.org/pipermail/gcc-patches/2023-November/638602.html
Which fixed the (new) tests baseptrs-[6789].C. I've noted that for now in
the patch, and adjusted the baseptrs-[46].C tests slightly to accommodate.
2024-01-08 Julian Brown <julian@codesourcery.com>
gcc/c-family/
* c-common.h (c_omp_address_inspector): Remove static from get_origin
and maybe_unconvert_ref methods.
* c-omp.cc (c_omp_split_clauses): Support OMP_ARRAY_SECTION.
(c_omp_address_inspector::map_supported_p): Handle OMP_ARRAY_SECTION.
(c_omp_address_inspector::get_origin): Avoid dereferencing possibly
NULL type when processing template decls.
(c_omp_address_inspector::maybe_unconvert_ref): Likewise.
gcc/cp/
* constexpr.cc (potential_consant_expression_1): Handle
OMP_ARRAY_SECTION.
* cp-tree.h (grok_omp_array_section, build_omp_array_section): Add
prototypes.
* decl2.cc (grok_omp_array_section): New function.
* error.cc (dump_expr): Handle OMP_ARRAY_SECTION.
* parser.cc (cp_parser_new): Initialize parser->omp_array_section_p.
(cp_parser_statement_expr): Disallow array sections.
(cp_parser_postfix_open_square_expression): Support OMP_ARRAY_SECTION
parsing.
(cp_parser_parenthesized_expression_list, cp_parser_lambda_expression,
cp_parser_braced_list): Disallow array sections.
(cp_parser_omp_var_list_no_open): Remove ALLOW_DEREF parameter, add
MAP_LVALUE in its place. Support generalised lvalue parsing for
OpenMP map, to and from clauses. Use OMP_ARRAY_SECTION
code instead of TREE_LIST to represent OpenMP array sections.
(cp_parser_omp_var_list): Remove ALLOW_DEREF parameter, add MAP_LVALUE.
Pass to cp_parser_omp_var_list_no_open.
(cp_parser_oacc_data_clause): Update call to cp_parser_omp_var_list.
(cp_parser_omp_clause_map): Add sk_omp scope around
cp_parser_omp_var_list_no_open call.
* parser.h (cp_parser): Add omp_array_section_p field.
* pt.cc (tsubst, tsubst_copy, tsubst_omp_clause_decl,
tsubst_copy_and_build): Add OMP_ARRAY_SECTION support.
* semantics.cc (handle_omp_array_sections_1, handle_omp_array_sections,
cp_oacc_check_attachments, finish_omp_clauses): Use OMP_ARRAY_SECTION
instead of TREE_LIST where appropriate. Handle more types of map
expression.
* typeck.cc (build_omp_array_section): New function.
gcc/
* gimplify.cc (gimplify_expr): Ensure OMP_ARRAY_SECTION has been
processed out before gimplification.
* tree-pretty-print.cc (dump_generic_node): Support OMP_ARRAY_SECTION.
* tree.def (OMP_ARRAY_SECTION): New tree code.
gcc/testsuite/
* c-c++-common/gomp/map-6.c: Update expected output.
* c-c++-common/gomp/target-enter-data-1.c: Update scan test.
* g++.dg/gomp/array-section-1.C: New test.
* g++.dg/gomp/array-section-2.C: New test.
* g++.dg/gomp/bad-array-section-1.C: New test.
* g++.dg/gomp/bad-array-section-2.C: New test.
* g++.dg/gomp/bad-array-section-3.C: New test.
* g++.dg/gomp/bad-array-section-4.C: New test.
* g++.dg/gomp/bad-array-section-5.C: New test.
* g++.dg/gomp/bad-array-section-6.C: New test.
* g++.dg/gomp/bad-array-section-7.C: New test.
* g++.dg/gomp/bad-array-section-8.C: New test.
* g++.dg/gomp/bad-array-section-9.C: New test.
* g++.dg/gomp/bad-array-section-10.C: New test.
* g++.dg/gomp/bad-array-section-11.C: New test.
* g++.dg/gomp/has_device_addr-non-lvalue-1.C: New test.
* g++.dg/gomp/pr67522.C: Update expected output.
* g++.dg/gomp/ind-base-3.C: New test.
* g++.dg/gomp/map-assignment-1.C: New test.
* g++.dg/gomp/map-inc-1.C: New test.
* g++.dg/gomp/map-lvalue-ref-1.C: New test.
* g++.dg/gomp/map-ptrmem-1.C: New test.
* g++.dg/gomp/map-ptrmem-2.C: New test.
* g++.dg/gomp/map-static-cast-lvalue-1.C: New test.
* g++.dg/gomp/map-ternary-1.C: New test.
* g++.dg/gomp/member-array-2.C: New test.
libgomp/
* testsuite/libgomp.c++/baseptrs-4.C: Remove commented-out cases that
now work.
* testsuite/libgomp.c++/baseptrs-6.C: New test.
* testsuite/libgomp.c++/ind-base-1.C: New test.
* testsuite/libgomp.c++/ind-base-2.C: New test.
* testsuite/libgomp.c++/lvalue-tofrom-1.C: New test.
* testsuite/libgomp.c++/lvalue-tofrom-2.C: New test.
* testsuite/libgomp.c++/map-comma-1.C: New test.
* testsuite/libgomp.c++/map-rvalue-ref-1.C: New test.
* testsuite/libgomp.c++/struct-ref-1.C: New test.
* testsuite/libgomp.c-c++-common/array-field-1.c: New test.
* testsuite/libgomp.c-c++-common/array-of-struct-1.c: New test.
* testsuite/libgomp.c-c++-common/array-of-struct-2.c: New test.
|
|
The following testcase ICEs during regimplificatgion since the addition of
(convert (eqne zero_one_valued_p@0 INTEGER_CST@1))
simplification. That simplification is novel in the sense that in
gimplify_expr it can turn an expression (comparison in particular) into
a SSA_NAME. Normally when gimplify_expr sees originally a SSA_NAME, it does
case SSA_NAME:
/* Allow callbacks into the gimplifier during optimization. */
ret = GS_ALL_DONE;
break;
and doesn't try to recalculate side effects because of that, but in this
case gimplify_expr normally enters the:
default:
switch (TREE_CODE_CLASS (TREE_CODE (*expr_p)))
{
case tcc_comparison:
then does
*expr_p = gimple_boolify (*expr_p);
and then
*expr_p = fold_convert_loc (input_location,
org_type, *expr_p);
with this new match.pd simplification turns that tcc_comparison class
into SSA_NAME. Unlike the outer SSA_NAME handling though, this falls
through into
recalculate_side_effects (*expr_p);
dont_recalculate:
break;
but unfortunately recalculate_side_effects doesn't handle SSA_NAME and ICEs
on it.
SSA_NAMEs don't ever have TREE_SIDE_EFFECTS set on those, so the following
patch fixes it by handling it similarly to the tcc_constant case.
2024-01-08 Jakub Jelinek <jakub@redhat.com>
PR tree-optimization/113228
* gimplify.cc (recalculate_side_effects): Do nothing for SSA_NAMEs.
* gcc.c-torture/compile/pr113228.c: New test.
|
|
|
|
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.
This version of the patch is based on the version posted for og13, and
additionally incorporates a follow-on fix for DECL_VALUE_EXPR handling
in gimplify_adjust_omp_clauses:
"OpenMP/OpenACC: Reorganise OMP map clause handling in gimplify.cc"
https://gcc.gnu.org/pipermail/gcc-patches/2023-June/622223.html
Parts of:
"OpenMP: OpenMP 5.2 semantics for pointers with unmapped target"
https://gcc.gnu.org/pipermail/gcc-patches/2023-June/623351.html
2023-12-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.
libgomp/
* testsuite/libgomp.fortran/target-enter-data-6.f90: Remove XFAIL.
|
|
This patch introduces enumerators to represent trait-set names and
trait names, which makes it easier to use tables to control other
behavior and for switch statements to dispatch on the tags. The tags
are stored in the same place in the TREE_LIST structure (OMP_TSS_ID or
OMP_TS_ID) and are encoded there as integer constants.
gcc/ChangeLog
* omp-selectors.h: New file.
* omp-general.h: Include omp-selectors.h.
(OMP_TSS_CODE, OMP_TSS_NAME): New.
(OMP_TS_CODE, OMP_TS_NAME): New.
(make_trait_set_selector, make_trait_selector): Adjust declarations.
(omp_construct_traits_to_codes): Likewise.
(omp_context_selector_set_compare): Likewise.
(omp_get_context_selector): Likewise.
(omp_get_context_selector_list): New.
* omp-general.cc (omp_construct_traits_to_codes): Pass length in
as argument instead of returning it. Make it table-driven.
(omp_tss_map): New.
(kind_properties, vendor_properties, extension_properties): New.
(atomic_default_mem_order_properties): New.
(omp_ts_map): New.
(omp_check_context_selector): Simplify lookup and dispatch logic.
(omp_mark_declare_variant): Ignore variants with unknown construct
selectors. Adjust for new representation.
(make_trait_set_selector, make_trait_selector): Adjust for new
representations.
(omp_context_selector_matches): Simplify dispatch logic. Avoid
fixed-sized buffers and adjust call to omp_construct_traits_to_codes.
(omp_context_selector_props_compare): Adjust for new representations
and simplify dispatch logic.
(omp_context_selector_set_compare): Likewise.
(omp_context_selector_compare): Likewise.
(omp_get_context_selector): Adjust for new representations, and split
out...
(omp_get_context_selector_list): New function.
(omp_lookup_tss_code): New.
(omp_lookup_ts_code): New.
(omp_context_compute_score): Adjust for new representations. Avoid
fixed-sized buffers and magic numbers. Adjust call to
omp_construct_traits_to_codes.
* gimplify.cc (omp_construct_selector_matches): Avoid use of
fixed-size buffer. Adjust call to omp_construct_traits_to_codes.
gcc/c/ChangeLog
* c-parser.cc (omp_construct_selectors): Delete.
(omp_device_selectors): Delete.
(omp_implementation_selectors): Delete.
(omp_user_selectors): Delete.
(c_parser_omp_context_selector): Adjust for new representations
and simplify dispatch logic. Uniformly warn instead of sometimes
error when an unknown selector is found. Adjust error messages
for extraneous/incorrect score.
(c_parser_omp_context_selector_specification): Likewise.
(c_finish_omp_declare_variant): Adjust for new representations.
gcc/cp/ChangeLog
* decl.cc (omp_declare_variant_finalize_one): Adjust for new
representations.
* parser.cc (omp_construct_selectors): Delete.
(omp_device_selectors): Delete.
(omp_implementation_selectors): Delete.
(omp_user_selectors): Delete.
(cp_parser_omp_context_selector): Adjust for new representations
and simplify dispatch logic. Uniformly warn instead of sometimes
error when an unknown selector is found. Adjust error messages
for extraneous/incorrect score.
(cp_parser_omp_context_selector_specification): Likewise.
* pt.cc (tsubst_attribute): Adjust for new representations.
gcc/fortran/ChangeLog
* gfortran.h: Include omp-selectors.h.
(enum gfc_omp_trait_property_kind): Delete, and replace all
references with equivalent omp_tp_type enumerators.
(struct gfc_omp_trait_property): Update for omp_tp_type.
(struct gfc_omp_selector): Replace string name with new enumerator.
(struct gfc_omp_set_selector): Likewise.
* openmp.cc (gfc_free_omp_trait_property_list): Update for
omp_tp_type.
(omp_construct_selectors): Delete.
(omp_device_selectors): Delete.
(omp_implementation_selectors): Delete.
(omp_user_selectors): Delete.
(gfc_ignore_trait_property_extension): New.
(gfc_ignore_trait_property_extension_list): New.
(gfc_match_omp_selector): Adjust for new representations and simplify
dispatch logic. Uniformly warn instead of sometimes error when an
unknown selector is found.
(gfc_match_omp_context_selector): Adjust for new representations.
Adjust error messages for extraneous/incorrect score.
(gfc_match_omp_context_selector_specification): Likewise.
* trans-openmp.cc (gfc_trans_omp_declare_variant): Adjust for
new representations.
gcc/testsuite/
* c-c++-common/gomp/declare-variant-1.c: Expect warning on
unknown selectors.
* c-c++-common/gomp/declare-variant-2.c: Likewise. Also adjust
messages for score errors.
* c-c++-common/gomp/declare-variant-no-score.c: New.
* gfortran.dg/gomp/declare-variant-1.f90: Expect warning on
unknown selectors.
* gfortran.dg/gomp/declare-variant-2.f90: Likewise. Also adjust
messages for score errors.
* gfortran.dg/gomp/declare-variant-no-score.f90: New.
|
|
This patch hides the underlying nested TREE_LIST structure of context
selectors behind accessor macros that have more meaningful names than
the generic TREE_PURPOSE/TREE_VALUE accessors. There is a slight
change to the representation in that the score expression in
trait-selectors has a distinguished tag and is separated from the
ordinary properties, although internally it is still represented as
the first item in the TREE_VALUE of the selector. This patch also renames
some local variables with slightly more descriptive names so it is easier
to track whether something is a selector-set, selector, or property.
gcc/ChangeLog
* omp-general.h (OMP_TS_SCORE_NODE): New.
(OMP_TSS_ID, OMP_TSS_TRAIT_SELECTORS): New.
(OMP_TS_ID, OMP_TS_SCORE, OMP_TS_PROPERTIES): New.
(OMP_TP_NAME, OMP_TP_VALUE): New.
(make_trait_set_selector): Declare.
(make_trait_selector): Declare.
(make_trait_property): Declare.
(omp_constructor_traits_to_codes): Rename to
omp_construct_traits_to_codes.
* omp-general.cc (omp_constructor_traits_to_codes): Rename
to omp_construct_traits_to_codes. Use new accessors.
(omp_check_context_selector): Use new accessors.
(make_trait_set_selector): New.
(make_trait_selector): New.
(make_trait_property): New.
(omp_context_name_list_prop): Use new accessors.
(omp_context_selector_matches): Use new accessors.
(omp_context_selector_props_compare): Use new accessors.
(omp_context_selector_set_compare): Use new accessors.
(omp_get_context_selector): Use new accessors.
(omp_context_compute_score): Use new accessors.
* gimplify.cc (omp_construct_selector_matches): Adjust for renaming
of omp_constructor_traits_to_codes.
gcc/c/ChangeLog
* c-parser.cc (c_parser_omp_context_selector): Use new constructors.
gcc/cp/ChangeLog
* parser.cc (cp_parser_omp_context_selector): Use new constructors.
* pt.cc: Include omp-general.h.
(tsubst_attribute): Use new context selector accessors and
constructors.
gcc/fortran/ChangeLog
* trans-openmp.cc (gfc_trans_omp_declare_variant): Use new
constructors.
|
|
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
(though that is not yet applied to mainline).
2023-08-18 Julian Brown <julian@codesourcery.com>
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.
|
|
This patch changes the mapping node arrangement used for array components
of derived types in order to accommodate for changes made in the previous
patch, particularly the use of "GOMP_MAP_ATTACH_DETACH" for pointer-typed
derived-type members instead of "GOMP_MAP_ALWAYS_POINTER".
We change the mapping nodes used for a derived-type mapping like this:
type T
integer, pointer, dimension(:) :: arrptr
end type T
type(T) :: tvar
[...]
!$omp target map(tofrom: tvar%arrptr)
So that the nodes used 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. A new
flag is introduced so the middle-end knows when the latter two kinds
are being used specifically for an array descriptor.
This version of the patch fixes "omp target exit data" handling
for GOMP_MAP_DELETE, and adds pretty-printing dump output
for the OMP_CLAUSE_RELEASE_DESCRIPTOR flag (for a little extra
clarity).
Also I noticed the handling of descriptors on *OpenACC*
exit-data directives was inconsistent, so I've made those use
GOMP_MAP_RELEASE/GOMP_MAP_DELETE with the new flag in the same way as
OpenMP too. In the end it doesn't actually matter to the runtime,
which handles GOMP_MAP_RELEASE/GOMP_MAP_DELETE/GOMP_MAP_TO_PSET for
array descriptors on OpenACC "exit data" directives the same, anyway,
and doing it this way in the FE avoids needless divergence.
I've added a couple of new tests (gomp/target-enter-exit-data.f90 and
goacc/enter-exit-data-2.f90).
2023-12-07 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.
* tree-pretty-print.cc (dump_omp_clause): Show
OMP_CLAUSE_RELEASE_DESCRIPTOR in dump output (with
GOMP_MAP_TO_PSET-like syntax).
gcc/testsuite/
* gfortran.dg/goacc/enter-exit-data-2.f90: New test.
* gfortran.dg/goacc/finalize-1.f: Adjust scan output.
* 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.
* gfortran.dg/gomp/target-enter-exit-data.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.
|
|
This patch reworks clause expansion in the C, C++ and (to a lesser
extent) Fortran front ends for OpenMP and OpenACC mapping nodes used in
GPU offloading support.
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 addresses several first-pass review comments
from Tobias, and fixes a few previously-missed cases for manually-managed
ragged array mappings (including cases using references). Some arbitrary
differences between handling of clause expansion for C vs. C++ have also
been fixed, and some fragments from later in the patch series have been
moved forward (where they were useful for fixing bugs). Several new
test cases have been added.
2023-11-29 Julian Brown <julian@codesourcery.com>
gcc/c-family/
* c-common.h (c_omp_region_type): Add C_ORT_EXIT_DATA,
C_ORT_OMP_EXIT_DATA and 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_P 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_data, c_parser_omp_target_enter_data): Support
ATTACH kind.
(c_parser_omp_target_exit_data): Support DETACH kind.
(check_clauses): Handle GOMP_MAP_POINTER and GOMP_MAP_ATTACH here.
* 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 "is OpenACC"
condition for C_ORT_ACC_TARGET addition.
gcc/cp/
* parser.cc (cp_parser_oacc_all_clauses): Add TARGET_P parameter. Use
to select region type for finish_omp_clauses call.
(cp_parser_omp_target_data, 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 OpenACC 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.
(finish_omp_target): Handle GOMP_MAP_POINTER.
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_get_attachment): Handle GOMP_MAP_DETACH here.
(omp_group_last): Handle GOMP_MAP_*, GOMP_MAP_DETACH,
GOMP_MAP_ATTACH_DETACH groups for "exit data" of reference-to-pointer
component array sections.
(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_first_chained_access_token): New
functions.
(omp_check_mapping_compatibility): Adjust accepted node combinations
for "from" clauses using release instead of alloc.
(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, tree-pretty-print.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.f90: Adjust scan output.
libgomp/
* target.c (gomp_map_pointer): Modify zero-length array section
pointer handling.
(gomp_attach_pointer): Likewise.
(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-c++-common/baseptrs-1.c: New test.
* testsuite/libgomp.c-c++-common/baseptrs-2.c: New test.
* testsuite/libgomp.c-c++-common/baseptrs-6.c: New test.
* testsuite/libgomp.c-c++-common/baseptrs-7.c: New test.
* testsuite/libgomp.c-c++-common/ptr-attach-2.c: New test.
* testsuite/libgomp.c-c++-common/target-implicit-map-2.c: Fix missing
"free".
* testsuite/libgomp.c-c++-common/target-implicit-map-5.c: New test.
* testsuite/libgomp.c-c++-common/target-map-zlas-1.c: New test.
* testsuite/libgomp.c++/class-array-1.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++/baseptrs-8.C: New test.
* testsuite/libgomp.c++/baseptrs-9.C: New test.
* testsuite/libgomp.c++/ref-mapping-1.C: New test.
* testsuite/libgomp.c++/target-48.C: New test.
* testsuite/libgomp.c++/target-49.C: New test.
* testsuite/libgomp.c++/target-exit-data-reftoptr-1.C: New test.
* testsuite/libgomp.c++/target-lambda-1.C: Update for OpenMP 5.2
semantics.
* testsuite/libgomp.c++/target-this-3.C: Likewise.
* testsuite/libgomp.c++/target-this-4.C: Likewise.
* testsuite/libgomp.fortran/struct-elem-map-1.f90: Add temporary XFAIL.
* testsuite/libgomp.fortran/target-enter-data-6.f90: Likewise.
|
|
In discussion of PR71093 it came up that more clobber_kind options would be
useful within the C++ front-end.
gcc/ChangeLog:
* tree-core.h (enum clobber_kind): Rename CLOBBER_EOL to
CLOBBER_STORAGE_END. Add CLOBBER_STORAGE_BEGIN,
CLOBBER_OBJECT_BEGIN, CLOBBER_OBJECT_END.
* gimple-lower-bitint.cc
* gimple-ssa-warn-access.cc
* gimplify.cc
* tree-inline.cc
* tree-ssa-ccp.cc: Adjust for rename.
* tree-pretty-print.cc: And handle new values.
gcc/cp/ChangeLog:
* call.cc (build_trivial_dtor_call): Use CLOBBER_OBJECT_END.
* decl.cc (build_clobber_this): Take clobber_kind argument.
(start_preparsed_function): Pass CLOBBER_OBJECT_BEGIN.
(begin_destructor_body): Pass CLOBBER_OBJECT_END.
gcc/testsuite/ChangeLog:
* gcc.dg/pr87052.c: Adjust expected CLOBBER output.
Co-authored-by: Nathaniel Shead <nathanieloshead@gmail.com>
|
|
The following avoids turning aggregate copy involving non-default
address-spaces to memcpy since that is not prepared for that.
GIMPLE verification no longer accepts WITH_SIZE_EXPR in aggregate
copies, the following re-allows that for the RHS. I also needed
to adjust one assert in DCE.
get_memory_address is used for string builtin expansion, so instead
of fixing that up for non-generic address-spaces I've put an assert
there.
I'll note that the same issue exists for initialization from an
empty CTOR which we gimplify to a memset call but since we are
not prepared to handle RTL expansion of the original VLA init and
I failed to provide test coverage (without extending the GNU C
extension for VLA structs) and the Ada frontend (or other frontends)
to not have address-space support the patch instead asserts we only
see generic address-spaces there.
PR middle-end/112830
* gimplify.cc (gimplify_modify_expr): Avoid turning aggregate
copy of non-generic address-spaces to memcpy.
(gimplify_modify_expr_to_memcpy): Assert we are dealing with
a copy inside the generic address-space.
(gimplify_modify_expr_to_memset): Likewise.
* tree-cfg.cc (verify_gimple_assign_single): Allow
WITH_SIZE_EXPR as part of the RHS of an assignment.
* builtins.cc (get_memory_address): Assert we are dealing
with the generic address-space.
* tree-ssa-dce.cc (ref_may_be_aliased): Handle WITH_SIZE_EXPR.
* gcc.target/avr/pr112830.c: New testcase.
* gcc.target/i386/pr112830.c: Likewise.
|
|
The new warning has two purposes: First, it makes clearer to the
user that it is about OpenMP and, secondly and more importantly,
it permits to use -Wno-openmp.
The newly added -Wopenmp is enabled by default and replaces the
'0' (always warning) in several OpenMP-related warning calls.
For code shared with OpenACC, it only uses OPT_Wopenmp for
'flag_openmp | flag_openmp_simd'.
gcc/c-family/ChangeLog:
* c.opt (Wopenmp): Add, enable by default.
gcc/c/ChangeLog:
* c-parser.cc (c_parser_omp_clause_num_threads,
c_parser_omp_clause_num_tasks, c_parser_omp_clause_grainsize,
c_parser_omp_clause_priority, c_parser_omp_clause_schedule,
c_parser_omp_clause_num_teams, c_parser_omp_clause_thread_limit,
c_parser_omp_clause_dist_schedule, c_parser_omp_depobj,
c_parser_omp_scan_loop_body, c_parser_omp_assumption_clauses):
Add OPT_Wopenmp to warning_at.
gcc/cp/ChangeLog:
* parser.cc (cp_parser_omp_clause_dist_schedule,
cp_parser_omp_scan_loop_body, cp_parser_omp_assumption_clauses,
cp_parser_omp_depobj): Add OPT_Wopenmp to warning_at.
* semantics.cc (finish_omp_clauses): Likewise.
gcc/ChangeLog:
* doc/invoke.texi (-Wopenmp): Add.
* gimplify.cc (gimplify_omp_for): Add OPT_Wopenmp to warning_at.
* omp-expand.cc (expand_omp_ordered_sink): Likewise.
* omp-general.cc (omp_check_context_selector): Likewise.
* omp-low.cc (scan_omp_for, check_omp_nesting_restrictions,
lower_omp_ordered_clauses): Likewise.
* omp-simd-clone.cc (simd_clone_clauses_extract): Likewise.
gcc/fortran/ChangeLog:
* lang.opt (Wopenmp): Add, enabled by dafault and documented in C.
* openmp.cc (gfc_match_omp_declare_target, resolve_positive_int_expr,
resolve_nonnegative_int_expr, resolve_omp_clauses,
gfc_resolve_omp_do_blocks): Use OPT_Wopenmp with gfc_warning{,_now}.
|
|
The following patch implements
CWG 2406 - [[fallthrough]] attribute and iteration statements
The genericization of some loops leaves nothing at all or just a label
after a body of a loop, so if the loop is later followed by
case or default label in a switch, the fallthrough statement isn't
diagnosed.
The following patch implements it by marking the IFN_FALLTHROUGH call
in such a case, such that during gimplification it can be pedantically
diagnosed even if it is followed by case or default label or some normal
labels followed by case/default labels.
While looking into this, I've discovered other problems.
expand_FALLTHROUGH_r is removing the IFN_FALLTHROUGH calls from the IL,
but wasn't telling that to walk_gimple_stmt/walk_gimple_seq_mod, so
the callers would then skip the next statement after it, and it would
return non-NULL if the removed stmt was last in the sequence. This could
lead to wi->callback_result being set even if it didn't appear at the very
end of switch sequence.
The patch makes use of wi->removed_stmt such that the callers properly
know what happened, and use different way to handle the end of switch
sequence case.
That change discovered a bug in the gimple-walk handling of
wi->removed_stmt. If that flag is set, the callback is telling the callers
that the current statement has been removed and so the innermost
walk_gimple_seq_mod shouldn't gsi_next. The problem is that
wi->removed_stmt is only reset at the start of a walk_gimple_stmt, but that
can be too late for some cases. If we have two nested gimple sequences,
say GIMPLE_BIND as the last stmt of some gimple seq, we remove the last
statement inside of that GIMPLE_BIND, set wi->removed_stmt there, don't
do gsi_next correctly because already gsi_remove moved us to the next stmt,
there is no next stmt, so we return back to the caller, but wi->removed_stmt
is still set and so we don't do gsi_next even in the outer sequence, despite
the GIMPLE_BIND (etc.) not being removed. That means we walk the
GIMPLE_BIND with its whole sequence again.
The patch fixes that by resetting wi->removed_stmt after we've used that
flag in walk_gimple_seq_mod. Nothing really uses that flag after the
outermost walk_gimple_seq_mod, it is just a private notification that
the stmt callback has removed a stmt.
2023-11-17 Jakub Jelinek <jakub@redhat.com>
PR c++/107571
gcc/
* gimplify.cc (expand_FALLTHROUGH_r): Use wi->removed_stmt after
gsi_remove, change the way of passing fallthrough stmt at the end
of sequence to expand_FALLTHROUGH. Diagnose IFN_FALLTHROUGH
with GF_CALL_NOTHROW flag.
(expand_FALLTHROUGH): Change loc into array of 2 location_t elts,
don't test wi.callback_result, instead check whether first
elt is not UNKNOWN_LOCATION and in that case pedwarn with the
second location.
* gimple-walk.cc (walk_gimple_seq_mod): Clear wi->removed_stmt
after the flag has been used.
* internal-fn.def (FALLTHROUGH): Mention in comment the special
meaning of the TREE_NOTHROW/GF_CALL_NOTHROW flag on the calls.
gcc/c-family/
* c-gimplify.cc (genericize_c_loop): For C++ mark IFN_FALLTHROUGH
call at the end of loop body as TREE_NOTHROW.
gcc/testsuite/
* g++.dg/DRs/dr2406.C: New test.
|
|
This patch implements the 'self' clause for compute constructs: parallel,
kernels, and serial. This clause conditionally uses the local device
(the host mult-core CPU) as the executing device of the compute region.
The actual implementation of the "local device" device type inside libgomp
(presumably using pthreads) is still not yet completed, so the libgomp
side is still implemented the exact same as host-fallback mode. (so as of now,
it essentially behaves like the 'if' clause with the condition inverted)
gcc/c/ChangeLog:
* c-parser.cc (c_parser_oacc_compute_clause_self): New function.
(c_parser_oacc_all_clauses): Add new 'bool compute_p = false'
parameter, add parsing of self clause when compute_p is true.
(OACC_KERNELS_CLAUSE_MASK): Add PRAGMA_OACC_CLAUSE_SELF.
(OACC_PARALLEL_CLAUSE_MASK): Likewise,
(OACC_SERIAL_CLAUSE_MASK): Likewise.
(c_parser_oacc_compute): Adjust call to c_parser_oacc_all_clauses to
set compute_p argument to true.
* c-typeck.cc (c_finish_omp_clauses): Add OMP_CLAUSE_SELF case.
gcc/cp/ChangeLog:
* parser.cc (cp_parser_oacc_compute_clause_self): New function.
(cp_parser_oacc_all_clauses): Add new 'bool compute_p = false'
parameter, add parsing of self clause when compute_p is true.
(OACC_KERNELS_CLAUSE_MASK): Add PRAGMA_OACC_CLAUSE_SELF.
(OACC_PARALLEL_CLAUSE_MASK): Likewise,
(OACC_SERIAL_CLAUSE_MASK): Likewise.
(cp_parser_oacc_compute): Adjust call to c_parser_oacc_all_clauses to
set compute_p argument to true.
* pt.cc (tsubst_omp_clauses): Add OMP_CLAUSE_SELF case.
* semantics.cc (c_finish_omp_clauses): Add OMP_CLAUSE_SELF case, merged
with OMP_CLAUSE_IF case.
gcc/fortran/ChangeLog:
* gfortran.h (typedef struct gfc_omp_clauses): Add self_expr field.
* openmp.cc (enum omp_mask2): Add OMP_CLAUSE_SELF.
(gfc_match_omp_clauses): Add handling for OMP_CLAUSE_SELF.
(OACC_PARALLEL_CLAUSES): Add OMP_CLAUSE_SELF.
(OACC_KERNELS_CLAUSES): Likewise.
(OACC_SERIAL_CLAUSES): Likewise.
(resolve_omp_clauses): Add handling for omp_clauses->self_expr.
* trans-openmp.cc (gfc_trans_omp_clauses): Add handling of
clauses->self_expr and building of OMP_CLAUSE_SELF tree clause.
(gfc_split_omp_clauses): Add handling of self_expr field copy.
gcc/ChangeLog:
* gimplify.cc (gimplify_scan_omp_clauses): Add OMP_CLAUSE_SELF case.
(gimplify_adjust_omp_clauses): Likewise.
* omp-expand.cc (expand_omp_target): Add OMP_CLAUSE_SELF expansion code,
* omp-low.cc (scan_sharing_clauses): Add OMP_CLAUSE_SELF case.
* tree-core.h (enum omp_clause_code): Add OMP_CLAUSE_SELF enum.
* tree-nested.cc (convert_nonlocal_omp_clauses): Add OMP_CLAUSE_SELF
case.
(convert_local_omp_clauses): Likewise.
* tree-pretty-print.cc (dump_omp_clause): Add OMP_CLAUSE_SELF case.
* tree.cc (omp_clause_num_ops): Add OMP_CLAUSE_SELF entry.
(omp_clause_code_name): Likewise.
* tree.h (OMP_CLAUSE_SELF_EXPR): New macro.
gcc/testsuite/ChangeLog:
* c-c++-common/goacc/self-clause-1.c: New test.
* c-c++-common/goacc/self-clause-2.c: New test.
* gfortran.dg/goacc/self.f95: New test.
include/ChangeLog:
* gomp-constants.h (GOACC_FLAG_LOCAL_DEVICE): New flag bit value.
libgomp/ChangeLog:
* oacc-parallel.c (GOACC_parallel_keyed): Add code to handle
GOACC_FLAG_LOCAL_DEVICE case.
* testsuite/libgomp.oacc-c-c++-common/self-1.c: New test.
|
|
gcc/ChangeLog:
* gimplify.cc (gimplify_bind_expr): Remove "omp allocate" attribute
to avoid that auxillary statement list reaches LTO.
gcc/testsuite/ChangeLog:
* gfortran.dg/gomp/allocate-13a.f90: New test.
|
|
gcc/fortran/ChangeLog:
* gfortran.h (ext_attr_t): Add omp_allocate flag.
* match.cc (gfc_free_omp_namelist): Void deleting same
u2.allocator multiple times now that a sequence can use
the same one.
* openmp.cc (gfc_match_omp_clauses, gfc_match_omp_allocate): Use
same allocator expr multiple times.
(is_predefined_allocator): Make static.
(gfc_resolve_omp_allocate): Update/extend restriction checks;
remove sorry message.
(resolve_omp_clauses): Reject corarrays in allocate/allocators
directive.
* parse.cc (check_omp_allocate_stmt): Permit procedure pointers
here (rejected later) for less misleading diagnostic.
* trans-array.cc (gfc_trans_auto_array_allocation): Propagate
size for GOMP_alloc and location to which it should be added to.
* trans-decl.cc (gfc_trans_deferred_vars): Handle 'omp allocate'
for stack variables; sorry for static variables/common blocks.
* trans-openmp.cc (gfc_trans_omp_clauses): Evaluate 'allocate'
clause's allocator only once; fix adding expressions to the
block.
(gfc_trans_omp_single): Pass a block to gfc_trans_omp_clauses.
gcc/ChangeLog:
* gimplify.cc (gimplify_bind_expr): Handle Fortran's
'omp allocate' for stack variables.
libgomp/ChangeLog:
* libgomp.texi (OpenMP Impl. Status): Mention that Fortran now
supports the allocate directive for stack variables.
* testsuite/libgomp.fortran/allocate-5.f90: New test.
* testsuite/libgomp.fortran/allocate-6.f90: New test.
* testsuite/libgomp.fortran/allocate-7.f90: New test.
* testsuite/libgomp.fortran/allocate-8.f90: New test.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/allocate-14.c: Fix directive name.
* c-c++-common/gomp/allocate-15.c: Likewise.
* c-c++-common/gomp/allocate-9.c: Fix comment typo.
* gfortran.dg/gomp/allocate-4.f90: Remove sorry dg-error.
* gfortran.dg/gomp/allocate-7.f90: Likewise.
* gfortran.dg/gomp/allocate-10.f90: New test.
* gfortran.dg/gomp/allocate-11.f90: New test.
* gfortran.dg/gomp/allocate-12.f90: New test.
* gfortran.dg/gomp/allocate-13.f90: New test.
* gfortran.dg/gomp/allocate-14.f90: New test.
* gfortran.dg/gomp/allocate-15.f90: New test.
* gfortran.dg/gomp/allocate-8.f90: New test.
* gfortran.dg/gomp/allocate-9.f90: New test.
|
|
Call GOMP_alloc/free for 'omp allocate' allocated variables. This is
for C only as C++ and Fortran show a sorry already in the FE. Note that
this only applies to stack variables as the C FE shows a sorry for
static variables.
gcc/ChangeLog:
* gimplify.cc (gimplify_bind_expr): Call GOMP_alloc/free for
'omp allocate' variables; move stack cleanup after other
cleanup.
(omp_notice_variable): Process original decl when decl
of the value-expression for a 'omp allocate' variable is passed.
* omp-low.cc (scan_omp_1_op): Handle 'omp allocate' variables
libgomp/ChangeLog:
* libgomp.texi (OpenMP 5.1 Impl.): Mark 'omp allocate' as
implemented for C only.
* testsuite/libgomp.c/allocate-4.c: New test.
* testsuite/libgomp.c/allocate-5.c: New test.
* testsuite/libgomp.c/allocate-6.c: New test.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/allocate-11.c: Remove C-only dg-message
for 'sorry, unimplemented'.
* c-c++-common/gomp/allocate-12.c: Likewise.
* c-c++-common/gomp/allocate-15.c: Likewise.
* c-c++-common/gomp/allocate-9.c: Likewise.
* c-c++-common/gomp/allocate-10.c: New test.
* c-c++-common/gomp/allocate-17.c: New test.
|
|
The 'allocate' directive can be used for both stack and static variables.
While the parser in C and C++ was pre-existing, it missed several
diagnostics, which this commit adds - for now only for C.
While the "sorry, unimplemented" for static variables is still issues
during parsing, the sorry for stack variables is now issued in the
middle end, preparing for the actual implementation. (Again: only for C.)
gcc/c/ChangeLog:
* c-parser.cc (c_parser_omp_construct): Move call to
c_parser_omp_allocate to ...
(c_parser_pragma): ... here.
(c_parser_omp_allocate): Avoid ICE is allocator could not be
parsed; set 'omp allocate' attribute for stack/automatic variables
and only reject static variables; add several additional
restriction checks.
* c-tree.h (c_mark_decl_jump_unsafe_in_current_scope): New prototype.
* c-decl.cc (decl_jump_unsafe): Return true for omp-allocated decls.
(c_mark_decl_jump_unsafe_in_current_scope): New.
(warn_about_goto, c_check_switch_jump_warnings): Add error for
omp-allocated decls.
gcc/ChangeLog:
* gimplify.cc (gimplify_bind_expr): Check for
insertion after variable cleanup. Convert 'omp allocate'
var-decl attribute to GOMP_alloc/GOMP_free calls.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/allocate-5.c: Fix testcase; make some
dg-messages for 'sorry' as c++, only.
* c-c++-common/gomp/directive-1.c: Make a 'sorry' c++ only.
* c-c++-common/gomp/allocate-9.c: New test.
* c-c++-common/gomp/allocate-11.c: New test.
* c-c++-common/gomp/allocate-12.c: New test.
* c-c++-common/gomp/allocate-14.c: New test.
* c-c++-common/gomp/allocate-15.c: New test.
* c-c++-common/gomp/allocate-16.c: New test.
|
|
In order to detect invalid jumps in and out of intervening code in
imperfectly-nested loops, the front ends need to insert some sort of
marker to identify the structured block sequences that they push into
the inner body of the loop. The error checking happens in the
diagnose_omp_blocks pass, between gimplification and OMP lowering, so
we need both GENERIC and GIMPLE representations of these markers.
They are removed in OMP lowering so no subsequent passes need to know
about them.
This patch doesn't include any front-end changes to generate the new
data structures.
gcc/cp/ChangeLog
* constexpr.cc (cxx_eval_constant_expression): Handle
OMP_STRUCTURED_BLOCK.
* pt.cc (tsubst_expr): Likewise.
gcc/ChangeLog
* doc/generic.texi (OpenMP): Document OMP_STRUCTURED_BLOCK.
* doc/gimple.texi (GIMPLE instruction set): Add
GIMPLE_OMP_STRUCTURED_BLOCK.
(GIMPLE_OMP_STRUCTURED_BLOCK): New subsection.
* gimple-low.cc (lower_stmt): Error on GIMPLE_OMP_STRUCTURED_BLOCK.
* gimple-pretty-print.cc (dump_gimple_omp_block): Handle
GIMPLE_OMP_STRUCTURED_BLOCK.
(pp_gimple_stmt_1): Likewise.
* gimple-walk.cc (walk_gimple_stmt): Likewise.
* gimple.cc (gimple_build_omp_structured_block): New.
* gimple.def (GIMPLE_OMP_STRUCTURED_BLOCK): New.
* gimple.h (gimple_build_omp_structured_block): Declare.
(gimple_has_substatements): Handle GIMPLE_OMP_STRUCTURED_BLOCK.
(CASE_GIMPLE_OMP): Likewise.
* gimplify.cc (is_gimple_stmt): Handle OMP_STRUCTURED_BLOCK.
(gimplify_expr): Likewise.
* omp-expand.cc (GIMPLE_OMP_STRUCTURED_BLOCK): Error on
GIMPLE_OMP_STRUCTURED_BLOCK.
* omp-low.cc (scan_omp_1_stmt): Handle GIMPLE_OMP_STRUCTURED_BLOCK.
(lower_omp_1): Likewise.
(diagnose_sb_1): Likewise.
(diagnose_sb_2): Likewise.
* tree-inline.cc (remap_gimple_stmt): Handle
GIMPLE_OMP_STRUCTURED_BLOCK.
(estimate_num_insns): Likewise.
* tree-nested.cc (convert_nonlocal_reference_stmt): Likewise.
(convert_local_reference_stmt): Likewise.
(convert_gimple_call): Likewise.
* tree-pretty-print.cc (dump_generic_node): Handle
OMP_STRUCTURED_BLOCK.
* tree.def (OMP_STRUCTURED_BLOCK): New.
* tree.h (OMP_STRUCTURED_BLOCK_BODY): New.
|
|
gcc/c-family/ChangeLog:
* c-format.cc (read_any_format_width):
Rename TRUE/FALSE to true/false.
gcc/ChangeLog:
* caller-save.cc (new_saved_hard_reg):
Rename TRUE/FALSE to true/false.
(setup_save_areas): Ditto.
* gcc.cc (set_collect_gcc_options): Ditto.
(driver::build_multilib_strings): Ditto.
(print_multilib_info): Ditto.
* genautomata.cc (gen_cpu_unit): Ditto.
(gen_query_cpu_unit): Ditto.
(gen_bypass): Ditto.
(gen_excl_set): Ditto.
(gen_presence_absence_set): Ditto.
(gen_presence_set): Ditto.
(gen_final_presence_set): Ditto.
(gen_absence_set): Ditto.
(gen_final_absence_set): Ditto.
(gen_automaton): Ditto.
(gen_regexp_repeat): Ditto.
(gen_regexp_allof): Ditto.
(gen_regexp_oneof): Ditto.
(gen_regexp_sequence): Ditto.
(process_decls): Ditto.
(reserv_sets_are_intersected): Ditto.
(initiate_excl_sets): Ditto.
(form_reserv_sets_list): Ditto.
(check_presence_pattern_sets): Ditto.
(check_absence_pattern_sets): Ditto.
(check_regexp_units_distribution): Ditto.
(check_unit_distributions_to_automata): Ditto.
(create_ainsns): Ditto.
(output_insn_code_cases): Ditto.
(output_internal_dead_lock_func): Ditto.
(form_important_insn_automata_lists): Ditto.
* gengtype-state.cc (read_state_files_list): Ditto.
* gengtype.cc (main): Ditto.
* gimple-array-bounds.cc (array_bounds_checker::check_array_bounds):
Ditto.
* gimple.cc (gimple_build_call_from_tree): Ditto.
(preprocess_case_label_vec_for_gimple): Ditto.
* gimplify.cc (gimplify_call_expr): Ditto.
* ordered-hash-map-tests.cc (test_map_of_int_to_strings): Ditto.
gcc/cp/ChangeLog:
* call.cc (build_conditional_expr):
Rename TRUE/FALSE to true/false.
(build_new_op): Ditto.
|
|
Both, specifying no category and specifying 'all', implies
that the implicit-behavior applies to all categories.
gcc/c/ChangeLog:
* c-parser.cc (c_parser_omp_clause_defaultmap): Parse
'all' as category.
gcc/cp/ChangeLog:
* parser.cc (cp_parser_omp_clause_defaultmap): Parse
'all' as category.
gcc/fortran/ChangeLog:
* gfortran.h (enum gfc_omp_defaultmap_category):
Add OMP_DEFAULTMAP_CAT_ALL.
* openmp.cc (gfc_match_omp_clauses): Parse
'all' as category.
* trans-openmp.cc (gfc_trans_omp_clauses): Handle it.
gcc/ChangeLog:
* tree-core.h (enum omp_clause_defaultmap_kind): Add
OMP_CLAUSE_DEFAULTMAP_CATEGORY_ALL.
* gimplify.cc (gimplify_scan_omp_clauses): Handle it.
* tree-pretty-print.cc (dump_omp_clause): Likewise.
libgomp/ChangeLog:
* libgomp.texi (OpenMP 5.2 status): Add depobj with
destroy-var argument as 'N'. Mark defaultmap with
'all' category as 'Y'.
gcc/testsuite/ChangeLog:
* gfortran.dg/gomp/defaultmap-1.f90: Update dg-error.
* c-c++-common/gomp/defaultmap-5.c: New test.
* c-c++-common/gomp/defaultmap-6.c: New test.
* gfortran.dg/gomp/defaultmap-10.f90: New test.
* gfortran.dg/gomp/defaultmap-9.f90: New test.
|
|
This patch implements the OpenACC 2.7 addition of default(none|present) support
for data constructs.
Now, specifying "default(none|present)" on a data construct turns on same
default clause behavior for all lexically enclosed compute constructs (which
don't already themselves have a default clause).
gcc/c/ChangeLog:
* c-parser.cc (OACC_DATA_CLAUSE_MASK): Add PRAGMA_OACC_CLAUSE_DEFAULT.
gcc/cp/ChangeLog:
* parser.cc (OACC_DATA_CLAUSE_MASK): Add PRAGMA_OACC_CLAUSE_DEFAULT.
gcc/fortran/ChangeLog:
* openmp.cc (OACC_DATA_CLAUSES): Add OMP_CLAUSE_DEFAULT.
gcc/ChangeLog:
* gimplify.cc (oacc_region_type_name): New function.
(oacc_default_clause): If no 'default' clause appears on this
compute construct, see if one appears on a lexically containing
'data' construct.
(gimplify_scan_omp_clauses): Upon OMP_CLAUSE_DEFAULT case, set
ctx->oacc_default_clause_ctx to current context.
gcc/testsuite/ChangeLog:
* c-c++-common/goacc/default-3.c: Adjust testcase.
* c-c++-common/goacc/default-4.c: Adjust testcase.
* c-c++-common/goacc/default-5.c: Adjust testcase.
* gfortran.dg/goacc/default-3.f95: Adjust testcase.
* gfortran.dg/goacc/default-4.f: Adjust testcase.
* gfortran.dg/goacc/default-5.f: Adjust testcase.
Co-authored-by: Thomas Schwinge <thomas@codesourcery.com>
|
|
This patch resolves PR c/110669, an ICE-after-error regression, by adding
a check that the array type isn't error_mark_node in gimplify_compound_lval.
2023-07-21 Roger Sayle <roger@nextmovesoftware.com>
Richard Biener <rguenther@suse.de>
gcc/ChangeLog
PR c/110699
* gimplify.cc (gimplify_compound_lval): If the array's type
is error_mark_node then return GS_ERROR.
gcc/testsuite/ChangeLog
PR c/110699
* gcc.dg/pr110699.c: New test case.
|
|
The manual references asm goto as being implicitly volatile already
and that was done when asm goto could not have outputs. When outputs
were added to `asm goto`, only asm goto without outputs were still being
marked as volatile. Now some parts of GCC decide, removing the `asm goto`
is ok if the output is not used, though not updating the CFG (this happens
on both the RTL level and the gimple level). Since the biggest user of `asm goto`
is the Linux kernel and they expect them to be volatile (they use them to
copy to/from userspace), we should just mark the inline-asm as volatile.
OK? Bootstrapped and tested on x86_64-linux-gnu.
PR middle-end/110420
PR middle-end/103979
PR middle-end/98619
gcc/ChangeLog:
* gimplify.cc (gimplify_asm_expr): Mark asm with labels as volatile.
gcc/testsuite/ChangeLog:
* gcc.c-torture/compile/asmgoto-6.c: New test.
|
|
Reduce number of enum values passed to libgomp as
GOMP_MAP_PRESENT_{TO,TOFROM,FROM,ALLOC} have the same semantic as
GOMP_MAP_FORCE_PRESENT (i.e. abort if not present, otherwise ignore);
that's different to GOMP_MAP_ALWAYS_PRESENT_{TO,TOFROM,FROM} which also
abort if not present but copy data when present. This is is a follow-up to
the commit r14-1579-g4ede915d5dde93 done 6 days ago.
Additionally, the commit improves a libgomp run-time and a C/C++ compile-time
error wording and extends testcases a tiny bit.
gcc/c/ChangeLog:
* c-parser.cc (c_parser_omp_clause_map): Reword error message for
clearness especially with 'omp target (enter/exit) data.'
gcc/cp/ChangeLog:
* parser.cc (cp_parser_omp_clause_map): Reword error message for
clearness especially with 'omp target (enter/exit) data.'
* semantics.cc (handle_omp_array_sections): Handle
GOMP_MAP_{ALWAYS_,}PRESENT_{TO,TOFROM,FROM,ALLOC} enum values.
gcc/ChangeLog:
* gimplify.cc (gimplify_adjust_omp_clauses_1): Use
GOMP_MAP_FORCE_PRESENT for 'present alloc' implicit mapping.
(gimplify_adjust_omp_clauses): Change
GOMP_MAP_PRESENT_{TO,TOFROM,FROM,ALLOC} to the equivalent
GOMP_MAP_FORCE_PRESENT.
* omp-low.cc (lower_omp_target): Remove handling of no-longer valid
GOMP_MAP_PRESENT_{TO,TOFROM,FROM,ALLOC}; update map kinds used for
to/from clauses with present modifier.
include/ChangeLog:
* gomp-constants.h (enum gomp_map_kind): Change the enum values
GOMP_MAP_PRESENT_{TO,TOFROM,FROM,ALLOC} to be compiler only.
(GOMP_MAP_PRESENT_P): Update to include also GOMP_MAP_FORCE_PRESENT.
libgomp/ChangeLog:
* target.c (gomp_to_device_kind_p, gomp_map_vars_internal): Replace
GOMP_MAP_PRESENT_{FROM,TO,TOFROM,ACLLOC} by GOMP_MAP_FORCE_PRESENT.
(gomp_map_vars_internal, gomp_update): Likewise; unify and improve
error message.
* testsuite/libgomp.c-c++-common/target-present-2.c: Update for
changed error message.
* testsuite/libgomp.fortran/target-present-1.f90: Likewise.
* testsuite/libgomp.fortran/target-present-2.f90: Likewise.
* testsuite/libgomp.oacc-c-c++-common/present-1.c: Likewise.
* testsuite/libgomp.c-c++-common/target-present-1.c: Likewise and
extend testcase to check that data is copied when needed.
* testsuite/libgomp.c-c++-common/target-present-3.c: Likewise.
* testsuite/libgomp.fortran/target-present-3.f90: Likewise.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/defaultmap-4.c: Update scan-tree-dump.
* c-c++-common/gomp/map-9.c: Likewise.
* gfortran.dg/gomp/defaultmap-8.f90: Likewise.
* gfortran.dg/gomp/map-11.f90: Likewise.
* gfortran.dg/gomp/target-update-1.f90: Likewise.
* gfortran.dg/gomp/map-12.f90: Likewise; also check original dump.
* c-c++-common/gomp/map-6.c: Update dg-error and also check
clause error with 'target (enter/exit) data'.
|
|
This implements support for the OpenMP 5.1 'present' modifier, which can be
used in map clauses in the 'target', 'target data', 'target data enter' and
'target data exit' constructs, and in the 'to' and 'from' clauses of the
'target update' construct. It is also supported in defaultmap.
The modifier triggers a fatal runtime error if the data specified by the
clause is not already present on the target device. It can also be combined
with 'always' in map clauses.
2023-06-06 Kwok Cheung Yeung <kcy@codesourcery.com>
Tobias Burnus <tobias@codesourcery.com>
gcc/c/
* c-parser.cc (c_parser_omp_clause_defaultmap,
c_parser_omp_clause_map): Parse 'present'.
(c_parser_omp_clause_to, c_parser_omp_clause_from): Remove.
(c_parser_omp_clause_from_to): New; parse to/from clauses with
optional present modifer.
(c_parser_omp_all_clauses): Update call.
(c_parser_omp_target_data, c_parser_omp_target_enter_data,
c_parser_omp_target_exit_data): Handle new map enum values
for 'present' mapping.
gcc/cp/
* parser.cc (cp_parser_omp_clause_defaultmap,
cp_parser_omp_clause_map): Parse 'present'.
(cp_parser_omp_clause_from_to): New; parse to/from
clauses with optional 'present' modifier.
(cp_parser_omp_all_clauses): Update call.
(cp_parser_omp_target_data, cp_parser_omp_target_enter_data,
cp_parser_omp_target_exit_data): Handle new enum value for
'present' mapping.
* semantics.cc (finish_omp_target): Likewise.
gcc/fortran/
* dump-parse-tree.cc (show_omp_namelist): Display 'present' map
modifier.
(show_omp_clauses): Display 'present' motion modifier for 'to'
and 'from' clauses.
* gfortran.h (enum gfc_omp_map_op): Add entries with 'present'
modifiers.
(struct gfc_omp_namelist): Add 'present_modifer'.
* openmp.cc (gfc_match_motion_var_list): New, handles optional
'present' modifier for to/from clauses.
(gfc_match_omp_clauses): Call it for to/from clauses; parse 'present'
in defaultmap and map clauses.
(resolve_omp_clauses): Allow 'present' modifiers on 'target',
'target data', 'target enter' and 'target exit' directives.
* trans-openmp.cc (gfc_trans_omp_clauses): Apply 'present' modifiers
to tree node for 'map', 'to' and 'from' clauses. Apply 'present' for
defaultmap.
gcc/
* gimplify.cc (omp_notice_variable): Apply GOVD_MAP_ALLOC_ONLY flag
and defaultmap flags if the defaultmap has GOVD_MAP_FORCE_PRESENT flag
set.
(omp_get_attachment): Handle map clauses with 'present' modifier.
(omp_group_base): Likewise.
(gimplify_scan_omp_clauses): Reorder present maps to come first.
Set GOVD flags for present defaultmaps.
(gimplify_adjust_omp_clauses_1): Set map kind for present defaultmaps.
* omp-low.cc (scan_sharing_clauses): Handle 'always, present' map
clauses.
(lower_omp_target): Handle map clauses with 'present' modifier.
Handle 'to' and 'from' clauses with 'present'.
* tree-core.h (enum omp_clause_defaultmap_kind): Add
OMP_CLAUSE_DEFAULTMAP_PRESENT defaultmap kind.
* tree-pretty-print.cc (dump_omp_clause): Handle 'map', 'to' and
'from' clauses with 'present' modifier. Handle present defaultmap.
* tree.h (OMP_CLAUSE_MOTION_PRESENT): New #define.
include/
* gomp-constants.h (GOMP_MAP_FLAG_SPECIAL_5): New.
(GOMP_MAP_FLAG_FORCE): Redefine.
(GOMP_MAP_FLAG_PRESENT, GOMP_MAP_FLAG_ALWAYS_PRESENT): New.
(enum gomp_map_kind): Add map kinds with 'present' modifiers.
(GOMP_MAP_COPY_TO_P, GOMP_MAP_COPY_FROM_P): Evaluate to true for
map variants with 'present'
(GOMP_MAP_ALWAYS_TO_P, GOMP_MAP_ALWAYS_FROM_P): Evaluate to true
for map variants with 'always, present' modifiers.
(GOMP_MAP_ALWAYS): Redefine.
(GOMP_MAP_FORCE_P, GOMP_MAP_PRESENT_P): New.
libgomp/
* libgomp.texi (OpenMP 5.1 Impl. status): Set 'present' support for
defaultmap to 'Y', add 'Y' entry for 'present' on to/from/map clauses.
* target.c (gomp_to_device_kind_p): Add map kinds with 'present'
modifier.
(gomp_map_vars_existing): Use new GOMP_MAP_FORCE_P macro.
(gomp_map_vars_internal, gomp_update, gomp_target_rev):
Emit runtime error if memory region not present.
* testsuite/libgomp.c-c++-common/target-present-1.c: New test.
* testsuite/libgomp.c-c++-common/target-present-2.c: New test.
* testsuite/libgomp.c-c++-common/target-present-3.c: New test.
* testsuite/libgomp.fortran/target-present-1.f90: New test.
* testsuite/libgomp.fortran/target-present-2.f90: New test.
* testsuite/libgomp.fortran/target-present-3.f90: New test.
gcc/testsuite/
* c-c++-common/gomp/map-6.c: Update dg-error, extend to test for
duplicated 'present' and extend scan-dump tests for 'present'.
* gfortran.dg/gomp/defaultmap-1.f90: Update dg-error.
* gfortran.dg/gomp/map-7.f90: Extend parse and dump test for
'present'.
* gfortran.dg/gomp/map-8.f90: Extend for duplicate 'present'
modifier checking.
* c-c++-common/gomp/defaultmap-4.c: New test.
* c-c++-common/gomp/map-9.c: New test.
* c-c++-common/gomp/target-update-1.c: New test.
* gfortran.dg/gomp/defaultmap-8.f90: New test.
* gfortran.dg/gomp/map-11.f90: New test.
* gfortran.dg/gomp/map-12.f90: New test.
* gfortran.dg/gomp/target-update-1.f90: New test.
|
|
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.
|
|
After the maybe_init_list_as_* patches, I noticed that we were putting the
array of strings into .rodata, but then memcpying it into an automatic
array, which is pointless; we should be able to use it directly.
This doesn't happen automatically because TREE_ADDRESSABLE is set (since
r12-657 for PR100464), and so gimplify_init_constructor won't promote the
variable to static. Theoretically we could do escape analysis to recognize
that the address, though taken, never leaves the function; that would allow
promotion when we're only using the address for indexing within the
function, as in initlist-opt2.C. But this would be a new pass.
And in initlist-opt1.C, we're passing the array address to another function,
so it definitely escapes; it's only safe in this case because it's calling a
standard library function that we know only uses it for indexing. So, a
flag seems needed. I first thought to put the flag on the TARGET_EXPR, but
the VAR_DECL seems more appropriate.
In a previous revision of the patch I called this flag DECL_NOT_OBSERVABLE,
but I think DECL_MERGEABLE is a better name, especially if we're going to
apply it to the backing array of initializer_list, which is observable. I
then also check it in places that check for -fmerge-all-constants, so that
multiple equivalent initializer-lists can also be combined. And then it
seemed to make sense for [[no_unique_address]] to have this meaning for
user-written variables.
I think the note in [dcl.init.list]/6 intended to allow this kind of merging
for initializer_lists, but it didn't actually work; for an explicit array
with the same initializer, if the address escapes the program could tell
whether the same variable in two frames have the same address. P2752 is
trying to correct this defect, so I'm going to assume that this is the
intent.
PR c++/110070
PR c++/105838
gcc/ChangeLog:
* tree.h (DECL_MERGEABLE): New.
* tree-core.h (struct tree_decl_common): Mention it.
* gimplify.cc (gimplify_init_constructor): Check it.
* cgraph.cc (symtab_node::address_can_be_compared_p): Likewise.
* varasm.cc (categorize_decl_for_section): Likewise.
gcc/cp/ChangeLog:
* call.cc (maybe_init_list_as_array): Set DECL_MERGEABLE.
(convert_like_internal) [ck_list]: Set it.
(set_up_extended_ref_temp): Copy it.
* tree.cc (handle_no_unique_addr_attribute): Set it.
gcc/testsuite/ChangeLog:
* g++.dg/tree-ssa/initlist-opt1.C: Check for static array.
* g++.dg/tree-ssa/initlist-opt2.C: Likewise.
* g++.dg/tree-ssa/initlist-opt4.C: New test.
* g++.dg/opt/icf1.C: New test.
* g++.dg/opt/icf2.C: New test.
* g++.dg/opt/icf3.C: New test.
* g++.dg/tree-ssa/array-temp1.C: Revert r12-657 change.
|
|
gcc/ChangeLog:
* alias.cc (ref_all_alias_ptr_type_p): Use _P() defines from tree.h.
* attribs.cc (diag_attr_exclusions): Ditto.
(decl_attributes): Ditto.
(build_type_attribute_qual_variant): Ditto.
* builtins.cc (fold_builtin_carg): Ditto.
(fold_builtin_next_arg): Ditto.
(do_mpc_arg2): Ditto.
* cfgexpand.cc (expand_return): Ditto.
* cgraph.h (decl_in_symtab_p): Ditto.
(symtab_node::get_create): Ditto.
* dwarf2out.cc (base_type_die): Ditto.
(implicit_ptr_descriptor): Ditto.
(gen_array_type_die): Ditto.
(gen_type_die_with_usage): Ditto.
(optimize_location_into_implicit_ptr): Ditto.
* expr.cc (do_store_flag): Ditto.
* fold-const.cc (negate_expr_p): Ditto.
(fold_negate_expr_1): Ditto.
(fold_convert_const): Ditto.
(fold_convert_loc): Ditto.
(constant_boolean_node): Ditto.
(fold_binary_op_with_conditional_arg): Ditto.
(build_fold_addr_expr_with_type_loc): Ditto.
(fold_comparison): Ditto.
(fold_checksum_tree): Ditto.
(tree_unary_nonnegative_warnv_p): Ditto.
(integer_valued_real_unary_p): Ditto.
(fold_read_from_constant_string): Ditto.
* gcc-rich-location.cc (maybe_range_label_for_tree_type_mismatch::get_text): Ditto.
* gimple-expr.cc (useless_type_conversion_p): Ditto.
(is_gimple_reg): Ditto.
(is_gimple_asm_val): Ditto.
(mark_addressable): Ditto.
* gimple-expr.h (is_gimple_variable): Ditto.
(virtual_operand_p): Ditto.
* gimple-ssa-warn-access.cc (pass_waccess::check_dangling_stores): Ditto.
* gimplify.cc (gimplify_bind_expr): Ditto.
(gimplify_return_expr): Ditto.
(gimple_add_padding_init_for_auto_var): Ditto.
(gimplify_addr_expr): Ditto.
(omp_add_variable): Ditto.
(omp_notice_variable): Ditto.
(omp_get_base_pointer): Ditto.
(omp_strip_components_and_deref): Ditto.
(omp_strip_indirections): Ditto.
(omp_accumulate_sibling_list): Ditto.
(omp_build_struct_sibling_lists): Ditto.
(gimplify_adjust_omp_clauses_1): Ditto.
(gimplify_adjust_omp_clauses): Ditto.
(gimplify_omp_for): Ditto.
(goa_lhs_expr_p): Ditto.
(gimplify_one_sizepos): Ditto.
* graphite-scop-detection.cc (scop_detection::graphite_can_represent_scev): Ditto.
* ipa-devirt.cc (odr_types_equivalent_p): Ditto.
* ipa-prop.cc (ipa_set_jf_constant): Ditto.
(propagate_controlled_uses): Ditto.
* ipa-sra.cc (type_prevails_p): Ditto.
(scan_expr_access): Ditto.
* optabs-tree.cc (optab_for_tree_code): Ditto.
* toplev.cc (wrapup_global_declaration_1): Ditto.
* trans-mem.cc (transaction_invariant_address_p): Ditto.
* tree-cfg.cc (verify_types_in_gimple_reference): Ditto.
(verify_gimple_comparison): Ditto.
(verify_gimple_assign_binary): Ditto.
(verify_gimple_assign_single): Ditto.
* tree-complex.cc (get_component_ssa_name): Ditto.
* tree-emutls.cc (lower_emutls_2): Ditto.
* tree-inline.cc (copy_tree_body_r): Ditto.
(estimate_move_cost): Ditto.
(copy_decl_for_dup_finish): Ditto.
* tree-nested.cc (convert_nonlocal_omp_clauses): Ditto.
(note_nonlocal_vla_type): Ditto.
(convert_local_omp_clauses): Ditto.
(remap_vla_decls): Ditto.
(fixup_vla_decls): Ditto.
* tree-parloops.cc (loop_has_vector_phi_nodes): Ditto.
* tree-pretty-print.cc (print_declaration): Ditto.
(print_call_name): Ditto.
* tree-sra.cc (compare_access_positions): Ditto.
* tree-ssa-alias.cc (compare_type_sizes): Ditto.
* tree-ssa-ccp.cc (get_default_value): Ditto.
* tree-ssa-coalesce.cc (populate_coalesce_list_for_outofssa): Ditto.
* tree-ssa-dom.cc (reduce_vector_comparison_to_scalar_comparison): Ditto.
* tree-ssa-forwprop.cc (can_propagate_from): Ditto.
* tree-ssa-propagate.cc (may_propagate_copy): Ditto.
* tree-ssa-sccvn.cc (fully_constant_vn_reference_p): Ditto.
* tree-ssa-sink.cc (statement_sink_location): Ditto.
* tree-ssa-structalias.cc (type_must_have_pointers): Ditto.
* tree-ssa-ter.cc (find_replaceable_in_bb): Ditto.
* tree-ssa-uninit.cc (warn_uninit): Ditto.
* tree-ssa.cc (maybe_rewrite_mem_ref_base): Ditto.
(non_rewritable_mem_ref_base): Ditto.
* tree-streamer-in.cc (lto_input_ts_type_non_common_tree_pointers): Ditto.
* tree-streamer-out.cc (write_ts_type_non_common_tree_pointers): Ditto.
* tree-vect-generic.cc (do_binop): Ditto.
(do_cond): Ditto.
* tree-vect-stmts.cc (vect_init_vector): Ditto.
* tree-vector-builder.h (tree_vector_builder::note_representative): Ditto.
* tree.cc (sign_mask_for): Ditto.
(verify_type_variant): Ditto.
(gimple_canonical_types_compatible_p): Ditto.
(verify_type): Ditto.
* ubsan.cc (get_ubsan_type_info_for_type): Ditto.
* var-tracking.cc (prepare_call_arguments): Ditto.
(vt_add_function_parameters): Ditto.
* varasm.cc (decode_addr_const): Ditto.
|
|
Previously, array descriptors might have been mapped as 'alloc'
instead of 'to' for 'alloc', not updating the array bounds. The
'alloc' could also appear for 'data exit', failing with a libgomp
assert. In some cases, either array descriptors or deferred-length
string's length variable was not mapped. And, finally, some offset
calculations with array-sections mappings went wrong.
Additionally, the patch now unmaps for scalar allocatables/pointers
the GOMP_MAP_POINTER, avoiding stale mappings.
The testcases contain some comment-out tests which require follow-up
work and for which PR exist. Those mostly relate to deferred-length
strings which have several issues beyong OpenMP support.
gcc/fortran/ChangeLog:
* trans-decl.cc (gfc_get_symbol_decl): Add attributes
such as 'declare target' also to hidden artificial
variable for deferred-length character variables.
* trans-openmp.cc (gfc_trans_omp_array_section,
gfc_trans_omp_clauses, gfc_trans_omp_target_exit_data):
Improve mapping of array descriptors and deferred-length
string variables.
gcc/ChangeLog:
* gimplify.cc (gimplify_scan_omp_clauses): Remove Fortran
special case.
libgomp/ChangeLog:
* testsuite/libgomp.fortran/target-enter-data-3.f90: Uncomment
'target exit data'.
* testsuite/libgomp.fortran/target-enter-data-4.f90: New test.
* testsuite/libgomp.fortran/target-enter-data-5.f90: New test.
* testsuite/libgomp.fortran/target-enter-data-6.f90: New test.
* testsuite/libgomp.fortran/target-enter-data-7.f90: New test.
gcc/testsuite/
* gfortran.dg/goacc/finalize-1.f: Update dg-tree; shows a fix
for 'finalize' as a ptr is now 'delete' instead of 'release'.
* gfortran.dg/gomp/pr78260-2.f90: Likewise as elem-size calc moved
to if (allocated) block
* gfortran.dg/gomp/target-exit-data.f90: Likewise as a var is now a
replaced by a MEM< _25 > expression.
* gfortran.dg/gomp/map-9.f90: Update dg-scan-tree-dump.
* gfortran.dg/gomp/map-10.f90: New test.
|
|
The Ada frontend, for example with gnat.dg/inline2_pkg.adb, tends
to create VIEW_CONVERT expressions with aggregate type even of
non-aggregate entities. In this case for example
return <retval> = (BIT_FIELD_REF <VIEW_CONVERT_EXPR<struct inline2_pkg__ieee_short_real>(number), 16, 16> & 32640) != 32640;
currently gimplification and SSA rewrite turn this into
_1 = BIT_FIELD_REF <VIEW_CONVERT_EXPR<struct inline2_pkg__ieee_short_real>(number_2(D));
which is two operations on a register. While as seen with PR109652
we might not want to completely rule out register to aggregate type
VIEW_CONVERTs we definitely do not want to stack multiple ops here.
The solution is to make sure the gimplifier puts a non-register as
the base object. For the above this will add
number.1 = number;
and use number.1 in the compound reference. Code generation is
unchanged, FRE optimizes this to BIT_FIELD_REF <number_2(D), ...>.
I think BIT_FIELD_REF <VIEW_CONVERT (x), ...> could be always
rewritten into BIT_FIELD_REF <x, ...>, but that's a separate thing.
* gimplify.cc (gimplify_compound_lval): When there's a
non-register type produced by one of the handled component
operations make sure we get a non-register base.
|
|
When for example complex lowering wants to extract the imaginary
part of a complex variable for lowering a complex move we can
end up with it generating __imag <VIEW_CONVERT_EXPR <_22> > which
is valid GENERIC. It then feeds that to the gimplifier via
force_gimple_operand but that fails to split up this chain
of handled components, generating invalid GIMPLE catched by
verification when PR109644 is fixed.
The following rectifies this by noting in gimplify_compound_lval
when the base object which we gimplify first ends up being a
register.
* gimplify.cc (gimplify_compound_lval): When the base
gimplified to a register make sure to split up chains
of operations.
|
|
On Wed, Feb 22, 2023 at 09:52:06AM +0000, Richard Biener wrote:
> > The following testcase ICEs because we still have some spots that
> > treat BUILT_IN_UNREACHABLE specially but not BUILT_IN_UNREACHABLE_TRAP
> > the same.
This patch uses (fndecl_built_in_p (node, BUILT_IN_UNREACHABLE)
|| fndecl_built_in_p (node, BUILT_IN_UNREACHABLE_TRAP))
a lot and from grepping around, we do something like that in lots of
other places, or in some spots instead as
(fndecl_built_in_p (node, BUILT_IN_NORMAL)
&& (DECL_FUNCTION_CODE (node) == BUILT_IN_WHATEVER1
|| DECL_FUNCTION_CODE (node) == BUILT_IN_WHATEVER2))
The following patch adds an overload for this case, so we can write
it in a shorter way, using C++11 argument packs so that it supports
as many codes as one needs.
2023-04-20 Jakub Jelinek <jakub@redhat.com>
Jonathan Wakely <jwakely@redhat.com>
* tree.h (built_in_function_equal_p): New helper function.
(fndecl_built_in_p): Turn into variadic template to support
1 or more built_in_function arguments.
* builtins.cc (fold_builtin_expect): Use 3 argument fndecl_built_in_p.
* gimplify.cc (goa_stabilize_expr): Likewise.
* cgraphclones.cc (cgraph_node::create_clone): Likewise.
* ipa-fnsummary.cc (compute_fn_summary): Likewise.
* omp-low.cc (setjmp_or_longjmp_p): Likewise.
* cgraph.cc (cgraph_edge::redirect_call_stmt_to_callee,
cgraph_update_edges_for_call_stmt_node,
cgraph_edge::verify_corresponds_to_fndecl,
cgraph_node::verify_node): Likewise.
* tree-stdarg.cc (optimize_va_list_gpr_fpr_size): Likewise.
* gimple-ssa-warn-access.cc (matching_alloc_calls_p): Likewise.
* ipa-prop.cc (try_make_edge_direct_virtual_call): Likewise.
|
|
With volatile, two 'x.data' comp refs aren't regarded as identical,
causing that the two items in the first map of
map(to:x.a, x.a.data) map(pset: x.a.data)
end up in separate 'map(struct:x)', which will cause a later ICE.
Solution: Ignore side effects when checking the operands in the hash
for being equal. (Do so by creating a variant of tree_operand_hash
that calls operand_equal_p with OEP_MATCH_SIDE_EFFECTS.)
gcc/ChangeLog:
PR middle-end/108545
* gimplify.cc (struct tree_operand_hash_no_se): New.
(omp_index_mapping_groups_1, omp_index_mapping_groups,
omp_reindex_mapping_groups, omp_mapped_by_containing_struct,
omp_tsort_mapping_groups_1, omp_tsort_mapping_groups,
oacc_resolve_clause_dependencies, omp_build_struct_sibling_lists,
gimplify_scan_omp_clauses): Use tree_operand_hash_no_se instead
of tree_operand_hash.
gcc/testsuite/ChangeLog:
PR middle-end/108545
* c-c++-common/gomp/map-8.c: New test.
* gfortran.dg/gomp/map-9.f90: New test.
|
|
This is a bug in tree-ssa-uninit.cc.
When doing the following:
/* Ignore the call to .DEFERRED_INIT that define the original
var itself as the following case:
temp = .DEFERRED_INIT (4, 2, “alt_reloc");
alt_reloc = temp;
In order to avoid generating warning for the fake usage
at alt_reloc = temp.
*/
We need to compare the var name inside the .DEFERRED_INIT call
(the 3rd argument) and the name for the LHS variable. if they are the same,
we will NOT report the warning.
There is one issue when we get the name for the LHS variable. when the
variable doesn't have a DECL_NAME (it's not a user declared variable,
which is the case for this bug):
_1 = .DEFERRED_INIT (4, 2, &"D.2389"[0]);
D.2389 = _1;
The current checking just ignores this case, and still report the warning.
The fix is very simple, when getting the name for the LHS variable, we should
consider this case and come up with the name the same way as we construct the
3rd argument for the call to .DEFERRED_INIT (please refer to the routine
"gimple_add_init_for_auto_var")
PR middle-end/107411
gcc/ChangeLog:
PR middle-end/107411
* gimplify.cc (gimple_add_init_for_auto_var): Use sprintf to replace
xasprintf.
* tree-ssa-uninit.cc (warn_uninit): Handle the case when the
LHS varaible of a .DEFERRED_INIT call doesn't have a DECL_NAME.
gcc/testsuite/ChangeLog:
PR middle-end/107411
* g++.dg/pr107411.C: New test.
|
|
gcc/
* gimplify.cc (gimplify_save_expr): Add missing guard.
gcc/ada/
* gcc-interface/trans.cc (gnat_gimplify_expr): Add missing guard.
gcc/testsuite/
* gnat.dg/shift2.adb: New test.
|
|
|
|
The following avoids passing down error_mark_node to fold_convert.
PR middle-end/107994
* gimplify.cc (gimplify_expr): Catch errorneous comparison
operand.
|
|
omp_{gs}et_teams_thread_limit on offload devices
This patch adds support for omp_get_max_teams, omp_set_num_teams, and
omp_{gs}et_teams_thread_limit on offload devices. That includes the usage of
device-specific ICV values (specified as environment variables or changed on a
device). In order to reuse device-specific ICV values, a copy back mechanism is
implemented that copies ICV values back from device to the host.
Additionally, a limitation of the number of teams on gcn offload devices is
implemented. The number of teams is limited by twice the number of compute
units (one team is executed on one compute unit). This avoids queueing
unnessecary many teams and a corresponding allocation of large amounts of
memory. Without that limitation the memory allocation for a large number of
user-specified teams can result in an "memory access fault".
A limitation of the number of teams is already also implemented for nvptx
devices (see nvptx_adjust_launch_bounds in libgomp/plugin/plugin-nvptx.c).
gcc/ChangeLog:
* gimplify.cc (optimize_target_teams): Set initial num_teams_upper
to "-2" instead of "1" for non-existing num_teams clause in order to
disambiguate from the case of an existing num_teams clause with value 1.
libgomp/ChangeLog:
* config/gcn/icv-device.c (omp_get_teams_thread_limit): Added to
allow processing of device-specific values.
(omp_set_teams_thread_limit): Likewise.
(ialias): Likewise.
* config/nvptx/icv-device.c (omp_get_teams_thread_limit): Likewise.
(omp_set_teams_thread_limit): Likewise.
(ialias): Likewise.
* icv-device.c (omp_get_teams_thread_limit): Likewise.
(ialias): Likewise.
(omp_set_teams_thread_limit): Likewise.
* icv.c (omp_set_teams_thread_limit): Removed.
(omp_get_teams_thread_limit): Likewise.
(ialias): Likewise.
* libgomp.texi: Updated documentation for nvptx and gcn corresponding
to the limitation of the number of teams.
* plugin/plugin-gcn.c (limit_teams): New helper function that limits
the number of teams by twice the number of compute units.
(parse_target_attributes): Limit the number of teams on gcn offload
devices.
* target.c (get_gomp_offload_icvs): Added teams_thread_limit_var
handling.
(gomp_load_image_to_device): Added a size check for the ICVs struct
variable.
(gomp_copy_back_icvs): New function that is used in GOMP_target_ext to
copy back the ICV values from device to host.
(GOMP_target_ext): Update the number of teams and threads in the kernel
args also considering device-specific values.
* testsuite/libgomp.c-c++-common/icv-4.c: Fixed an error in the reading
of OMP_TEAMS_THREAD_LIMIT from the environment.
* testsuite/libgomp.c-c++-common/icv-5.c: Extended.
* testsuite/libgomp.c-c++-common/icv-6.c: Extended.
* testsuite/libgomp.c-c++-common/icv-7.c: Extended.
* testsuite/libgomp.c-c++-common/icv-9.c: New test.
* testsuite/libgomp.fortran/icv-5.f90: New test.
* testsuite/libgomp.fortran/icv-6.f90: New test.
gcc/testsuite/ChangeLog:
* c-c++-common/gomp/target-teams-1.c: Adapt expected values for
num_teams from "1" to "-2" in cases without num_teams clause.
* g++.dg/gomp/target-teams-1.C: Likewise.
* gfortran.dg/gomp/defaultmap-4.f90: Likewise.
* gfortran.dg/gomp/defaultmap-5.f90: Likewise.
* gfortran.dg/gomp/defaultmap-6.f90: Likewise.
|
|
This another one of these ICE after error issues with the
gimplifier and a fallout from r12-3278-g823685221de986af.
The problem here is gimplify_modify_expr does not
check if either from or to was an error operand.
This adds the check and fixes the ICE.
OK? Bootstrapped and tested on x86_64-linux-gnu with no regressions.
gcc/ChangeLog:
* gimplify.cc (gimplify_modify_expr): If
either *from_p or *to_p were error_operand
return early.
gcc/testsuite/ChangeLog:
* gcc.dg/redecl-23.c: New test.
* gcc.dg/redecl-24.c: New test.
* gcc.dg/redecl-25.c: New test.
|
|
The problem here is the gimplifier returns GS_ERROR but
in some cases we don't check that soon enough and try
to do other work which could crash.
So the fix in these two cases is to return GS_ERROR
early if the gimplify_* functions had return GS_ERROR.
OK? Bootstrapped and tested on x86_64-linux-gnu with no regressions.
Thanks,
Andrew Pinski
gcc/ChangeLog:
PR c/106764
PR c/106765
PR c/107307
* gimplify.cc (gimplify_compound_lval): Return GS_ERROR
if gimplify_expr had return GS_ERROR.
(gimplify_call_expr): Likewise.
gcc/testsuite/ChangeLog:
PR c/106764
PR c/106765
PR c/107307
* gcc.dg/redecl-19.c: New test.
* gcc.dg/redecl-20.c: New test.
* gcc.dg/redecl-21.c: New test.
|
|
While looking at gimple_boolify for PR107368, I've noticed 2 comment
typos.
2022-10-25 Jakub Jelinek <jakub@redhat.com>
* gimplify.cc (gimple_boolify): Fix comment typos, prduce -> produce
and There -> These.
|
|
The following testcase ICEs in C, because assume attribute condition
has int type rather than bool and the gimplification into GIMPLE_ASSUME
assigns it into a bool variable.
Fixed by calling gimple_boolify.
2022-10-25 Jakub Jelinek <jakub@redhat.com>
PR tree-optimization/107368
* gimplify.cc (gimplify_call_expr): For complex IFN_ASSUME
conditions call gimple_boolify on the condition.
* gcc.dg/attr-assume-5.c: New test.
|
|
The FEs emit errors about jumps into assume attribute conditions,
but when we add GIMPLE_ASSUME for the condition which is reachable
through those jumps, we can run into cfg verification diagnostics.
Fixed by throwing the IFN_ASSUME away during gimplification if
seen_error () - like we already do for -O0. GIMPLE_ASSUME in the middle-end
is a pure optimization thing and if errors were reported, the optimizations
will not be beneficial for anything.
2022-10-25 Jakub Jelinek <jakub@redhat.com>
PR tree-optimization/107369
* gimplify.cc (gimplify_call_expr): If seen_error, handle complex
IFN_ASSUME the same as for -O0.
* gcc.dg/attr-assume-4.c: New test.
* g++.dg/cpp23/attr-assume8.C: New test.
|
|
Simplify several calls to build_string_literal by not requiring redundant
strlen or IDENTIFIER_* in the caller.
I also corrected a wrong comment on IDENTIFIER_LENGTH.
gcc/ChangeLog:
* tree.h (build_string_literal): New one-argument overloads that
take tree (identifier) and const char *.
* builtins.cc (fold_builtin_FILE)
(fold_builtin_FUNCTION)
* gimplify.cc (gimple_add_init_for_auto_var)
* vtable-verify.cc (verify_bb_vtables): Simplify calls.
gcc/cp/ChangeLog:
* cp-gimplify.cc (fold_builtin_source_location)
* vtable-class-hierarchy.cc (register_all_pairs): Simplify calls to
build_string_literal.
(build_string_from_id): Remove.
|
|
My earlier patches gimplify the simplest non-side-effects assumptions
into if (cond) ; else __builtin_unreachable (); and throw the rest
on the floor.
The following patch attempts to do something with the rest too.
For -O0, it throws the more complex assumptions on the floor,
we don't expect optimizations and the assumptions are there to allow
optimizations. Otherwise arranges for the assumptions to be
visible in the IL as
.ASSUME (_Z2f4i._assume.0, i_1(D));
call where there is an artificial function like:
bool _Z2f4i._assume.0 (int i)
{
bool _2;
<bb 2> [local count: 1073741824]:
_2 = i_1(D) == 43;
return _2;
}
with the semantics that there is UB unless the assumption function
would return true.
Aldy, could ranger handle this? If it sees .ASSUME call,
walk the body of such function from the edge(s) to exit with the
assumption that the function returns true, so above set _2 [true, true]
and from there derive that i_1(D) [43, 43] and then map the argument
in the assumption function to argument passed to IFN_ASSUME (note,
args there are shifted by 1)?
During gimplification it actually gimplifies it into
[[assume (D.2591)]]
{
{
i = i + 1;
D.2591 = i == 44;
}
}
which is a new GIMPLE_ASSUME statement wrapping a GIMPLE_BIND and
specifying a boolean_type_node variable which contains the result.
The GIMPLE_ASSUME then survives just a couple of passes and is lowered
during gimple lowering into an outlined separate function and
IFN_ASSUME call. Variables declared inside of the
condition (both static and automatic) just change context, automatic
variables from the caller are turned into parameters (note, as the code
is never executed, I handle this way even non-POD types, we don't need to
bother pretending there would be user copy constructors etc. involved).
The assume_function artificial functions are then optimized until the
new assumptions pass which doesn't do much right now but I'd like to see
there the backwards ranger walk and filling up of SSA_NAME_RANGE_INFO
for the parameters.
There are a few further changes I'd like to do, like ignoring the
.ASSUME calls in inlining size estimations (but haven't figured out where
it is done), or for LTO arrange for the assume functions to be emitted
in all partitions that reference those (usually there will be just one,
unless code with the assumption got inlined, versioned etc.).
2022-10-18 Jakub Jelinek <jakub@redhat.com>
PR c++/106654
gcc/
* gimple.def (GIMPLE_ASSUME): New statement kind.
* gimple.h (struct gimple_statement_assume): New type.
(is_a_helper <gimple_statement_assume *>::test,
is_a_helper <const gimple_statement_assume *>::test): New.
(gimple_build_assume): Declare.
(gimple_has_substatements): Return true for GIMPLE_ASSUME.
(gimple_assume_guard, gimple_assume_set_guard,
gimple_assume_guard_ptr, gimple_assume_body_ptr, gimple_assume_body):
New inline functions.
* gsstruct.def (GSS_ASSUME): New.
* gimple.cc (gimple_build_assume): New function.
(gimple_copy): Handle GIMPLE_ASSUME.
* gimple-pretty-print.cc (dump_gimple_assume): New function.
(pp_gimple_stmt_1): Handle GIMPLE_ASSUME.
* gimple-walk.cc (walk_gimple_op): Handle GIMPLE_ASSUME.
* omp-low.cc (WALK_SUBSTMTS): Likewise.
(lower_omp_1): Likewise.
* omp-oacc-kernels-decompose.cc (adjust_region_code_walk_stmt_fn):
Likewise.
* tree-cfg.cc (verify_gimple_stmt, verify_gimple_in_seq_2): Likewise.
* function.h (struct function): Add assume_function bitfield.
* gimplify.cc (gimplify_call_expr): If the assumption isn't
simple enough, expand it into GIMPLE_ASSUME wrapped block or
for -O0 drop it.
* gimple-low.cc: Include attribs.h.
(create_assumption_fn): New function.
(struct lower_assumption_data): New type.
(find_assumption_locals_r, assumption_copy_decl,
adjust_assumption_stmt_r, adjust_assumption_stmt_op,
lower_assumption): New functions.
(lower_stmt): Handle GIMPLE_ASSUME.
* tree-ssa-ccp.cc (pass_fold_builtins::execute): Remove
IFN_ASSUME calls.
* lto-streamer-out.cc (output_struct_function_base): Pack
assume_function bit.
* lto-streamer-in.cc (input_struct_function_base): And unpack it.
* cgraphunit.cc (cgraph_node::expand): Don't verify assume_function
has TREE_ASM_WRITTEN set and don't release its body.
(symbol_table::compile): Allow assume functions not to have released
body.
* internal-fn.cc (expand_ASSUME): Remove gcc_unreachable.
* passes.cc (execute_one_pass): For TODO_discard_function don't
release body of assume functions.
* cgraph.cc (cgraph_node::verify_node): Don't verify cgraph nodes
of PROP_assumptions_done functions.
* tree-pass.h (PROP_assumptions_done): Define.
(TODO_discard_function): Adjust comment.
(make_pass_assumptions): Declare.
* passes.def (pass_assumptions): Add.
* timevar.def (TV_TREE_ASSUMPTIONS): New.
* tree-inline.cc (remap_gimple_stmt): Handle GIMPLE_ASSUME.
* tree-vrp.cc (pass_data_assumptions): New variable.
(pass_assumptions): New class.
(make_pass_assumptions): New function.
gcc/cp/
* cp-tree.h (build_assume_call): Declare.
* parser.cc (cp_parser_omp_assumption_clauses): Use build_assume_call.
* cp-gimplify.cc (build_assume_call): New function.
(process_stmt_assume_attribute): Use build_assume_call.
* pt.cc (tsubst_copy_and_build): Likewise.
gcc/testsuite/
* g++.dg/cpp23/attr-assume5.C: New test.
* g++.dg/cpp23/attr-assume6.C: New test.
* g++.dg/cpp23/attr-assume7.C: New test.
|
|
In this testcase, we were optimizing away the temporary for f(), but
C++17 and above are clear that there is a temporary, and because its
destructor has visible side-effects we can't optimize it away under the
as-if rule. So disable this optimization for TREE_ADDRESSABLE type.
I moved the declaration of volatile_p after the call to
gimple_fold_indirect_ref_rhs to minimize indentation changes; I don't see
any way the value of that flag could be affected by the call.
gcc/ChangeLog:
* gimplify.cc (gimplify_modify_expr_rhs): Don't optimize
x = *(A*)&<expr> to x = <expr> for a TREE_ADDRESSABLE type.
gcc/testsuite/ChangeLog:
* g++.dg/init/elide9.C: New test.
|
|
The following patch implements C++23 P1774R8 - Portable assumptions
paper, by introducing support for [[assume (cond)]]; attribute for C++.
In addition to that the patch adds [[gnu::assume (cond)]]; and
__attribute__((assume (cond))); support to both C and C++.
As described in C++23, the attribute argument is conditional-expression
rather than the usual assignment-expression for attribute arguments,
the condition is contextually converted to bool (for C truthvalue conversion
is done on it) and is never evaluated at runtime.
For C++ constant expression evaluation, I only check the simplest conditions
for undefined behavior, because otherwise I'd need to undo changes to
*ctx->global which happened during the evaluation (but I believe the spec
allows that and we can further improve later).
The patch uses a new internal function, .ASSUME, to hold the condition
in the FEs. At gimplification time, if the condition is simple/without
side-effects, it is gimplified as if (cond) ; else __builtin_unreachable ();
and otherwise for now dropped on the floor. The intent is to incrementally
outline the conditions into separate artificial functions and use
.ASSUME further to tell the ranger and perhaps other optimization passes
about the assumptions, as detailed in the PR.
When implementing it, I found that assume entry hasn't been added to
https://eel.is/c++draft/cpp.cond#6
Jonathan said he'll file a NB comment about it, this patch assumes it
has been added into the table as 202207L when the paper has been voted in.
With the attributes for both C/C++, I'd say we don't need to add
__builtin_assume with similar purpose, especially when __builtin_assume
in LLVM is just weird. It is strange for side-effects in function call's
argument not to be evaluated, and LLVM in that case (annoyingly) warns
and ignores the side-effects (but doesn't do then anything with it),
if there are no side-effects, it will work like our
if (!cond) __builtin_unreachable ();
2022-10-06 Jakub Jelinek <jakub@redhat.com>
PR c++/106654
gcc/
* internal-fn.def (ASSUME): New internal function.
* internal-fn.h (expand_ASSUME): Declare.
* internal-fn.cc (expand_ASSUME): Define.
* gimplify.cc (gimplify_call_expr): Gimplify IFN_ASSUME.
* fold-const.h (simple_condition_p): Declare.
* fold-const.cc (simple_operand_p_2): Rename to ...
(simple_condition_p): ... this. Remove forward declaration.
No longer static. Adjust function comment and fix a typo in it.
Adjust recursive call.
(simple_operand_p): Adjust function comment.
(fold_truth_andor): Adjust simple_operand_p_2 callers to call
simple_condition_p.
* doc/extend.texi: Document assume attribute. Move fallthrough
attribute example to its section.
gcc/c-family/
* c-attribs.cc (handle_assume_attribute): New function.
(c_common_attribute_table): Add entry for assume attribute.
* c-lex.cc (c_common_has_attribute): Handle
__have_cpp_attribute (assume).
gcc/c/
* c-parser.cc (handle_assume_attribute): New function.
(c_parser_declaration_or_fndef): Handle assume attribute.
(c_parser_attribute_arguments): Add assume_attr argument,
if true, parse first argument as conditional expression.
(c_parser_gnu_attribute, c_parser_std_attribute): Adjust
c_parser_attribute_arguments callers.
(c_parser_statement_after_labels) <case RID_ATTRIBUTE>: Handle
assume attribute.
gcc/cp/
* cp-tree.h (process_stmt_assume_attribute): Implement C++23
P1774R8 - Portable assumptions. Declare.
(diagnose_failing_condition): Declare.
(find_failing_clause): Likewise.
* parser.cc (assume_attr): New enumerator.
(cp_parser_parenthesized_expression_list): Handle assume_attr.
Remove identifier variable, for id_attr push the identifier into
expression_list right away instead of inserting it before all the
others at the end.
(cp_parser_conditional_expression): New function.
(cp_parser_constant_expression): Use it.
(cp_parser_statement): Handle assume attribute.
(cp_parser_expression_statement): Likewise.
(cp_parser_gnu_attribute_list): Use assume_attr for assume
attribute.
(cp_parser_std_attribute): Likewise. Handle standard assume
attribute like gnu::assume.
* cp-gimplify.cc (process_stmt_assume_attribute): New function.
* constexpr.cc: Include fold-const.h.
(find_failing_clause_r, find_failing_clause): New functions,
moved from semantics.cc with ctx argument added and if non-NULL,
call cxx_eval_constant_expression rather than fold_non_dependent_expr.
(cxx_eval_internal_function): Handle IFN_ASSUME.
(potential_constant_expression_1): Likewise.
* pt.cc (tsubst_copy_and_build): Likewise.
* semantics.cc (diagnose_failing_condition): New function.
(find_failing_clause_r, find_failing_clause): Moved to constexpr.cc.
(finish_static_assert): Use it. Add auto_diagnostic_group.
gcc/testsuite/
* gcc.dg/attr-assume-1.c: New test.
* gcc.dg/attr-assume-2.c: New test.
* gcc.dg/attr-assume-3.c: New test.
* g++.dg/cpp2a/feat-cxx2a.C: Add colon to C++20 features
comment, add C++20 attributes comment and move C++20
new features after the attributes before them.
* g++.dg/cpp23/feat-cxx2b.C: Likewise. Test
__has_cpp_attribute(assume).
* g++.dg/cpp23/attr-assume1.C: New test.
* g++.dg/cpp23/attr-assume2.C: New test.
* g++.dg/cpp23/attr-assume3.C: New test.
* g++.dg/cpp23/attr-assume4.C: New test.
|
|
This patch is a minimal fix for the recently-added
struct-component-kind-1.c test (which is currently failing to emit one
of the errors it expects in scan output). This fragment was erroneously
omitted from the second version of the patch posted previously:
https://gcc.gnu.org/pipermail/gcc-patches/2022-September/602504.html
2022-10-01 Julian Brown <julian@codesourcery.com>
gcc/
* gimplify.cc (omp_group_base): Fix IF_PRESENT (no_create)
handling.
|
|
This patch fixes an ICE when both a complete struct variable and
components of that struct are mapped on the same directive for OpenACC,
using a modified version of the scheme used for OpenMP in the following
patch:
https://gcc.gnu.org/pipermail/gcc-patches/2022-September/601558.html
A new function has been added to make sure that the mapping kinds of
the whole struct and the member access are compatible -- conservatively,
so as not to copy more to/from the device than the user expects.
This version of the patch uses a different method to detect duplicate
clauses for OpenACC in oacc_resolve_clause_dependencies, and removes
the now-redundant check in omp_accumulate_sibling_lists. (The latter
check would no longer trigger when we map the whole struct on the same
directive because the component-mapping clauses are now deleted before
the check is executed.)
2022-09-28 Julian Brown <julian@codesourcery.com>
gcc/
PR middle-end/107028
* gimplify.cc (omp_check_mapping_compatibility,
oacc_resolve_clause_dependencies): New functions.
(omp_accumulate_sibling_list): Remove redundant duplicate clause
detection for OpenACC.
(build_struct_sibling_lists): Skip deleted groups. Don't build sibling
list for struct variables that are fully mapped on the same directive
for OpenACC.
(gimplify_scan_omp_clauses): Call oacc_resolve_clause_dependencies.
gcc/testsuite/
PR middle-end/107028
* c-c++-common/goacc/struct-component-kind-1.c: New test.
* g++.dg/goacc/pr107028-1.C: New test.
* g++.dg/goacc/pr107028-2.C: New test.
* gfortran.dg/goacc/mapping-tests-5.f90: New test.
|
