From 9a668532fb19e7c57aa595a26ce3f0d95f9cbb1b Mon Sep 17 00:00:00 2001 From: Tobias Burnus Date: Fri, 1 Jul 2022 17:52:03 +0200 Subject: OpenMP: Handle tofrom with target enter/exit data In 5.2, a map clause can be map-entering or map-exiting, either containing 'tofrom'. The main reason for this is permit 'map(x)' with 'omp target enter/exit data', avoiding to specify 'to:/from:' explicitly. (OpenMP defaults to 'tofrom'.) gcc/c/ChangeLog: * c-parser.cc (c_parser_omp_target_enter_data, c_parser_omp_target_exit_data): Accept tofrom map-type modifier but use 'to' / 'from' internally. gcc/cp/ChangeLog: * parser.cc (cp_parser_omp_target_enter_data, cp_parser_omp_target_exit_data): Accept tofrom map-type modifier but use 'to' / 'from' internally. gcc/fortran/ChangeLog: * dump-parse-tree.cc (show_omp_namelist): For the map-type, also handle the always modifer and release/delete. * openmp.cc (resolve_omp_clauses): Accept tofrom map-type modifier for target enter/exit data, but use 'to' / 'from' internally. libgomp/ChangeLog: * libgomp.texi (OpenMP 5.2): Mark target enter/exit data with fromto as implemented. gcc/testsuite/ChangeLog: * c-c++-common/gomp/target-data-2.c: New test. * c-c++-common/gomp/target-data-3.c: New test. * gfortran.dg/gomp/target-data-1.f90: New test. * gfortran.dg/gomp/target-data-2.f90: New test. --- gcc/c/c-parser.cc | 22 +++++++++++++++++++--- 1 file changed, 19 insertions(+), 3 deletions(-) (limited to 'gcc/c') diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc index 1704a52..97e3b23 100644 --- a/gcc/c/c-parser.cc +++ b/gcc/c/c-parser.cc @@ -21072,6 +21072,14 @@ c_parser_omp_target_enter_data (location_t loc, c_parser *parser, case GOMP_MAP_ALLOC: map_seen = 3; break; + case GOMP_MAP_TOFROM: + OMP_CLAUSE_SET_MAP_KIND (*pc, GOMP_MAP_TO); + map_seen = 3; + break; + case GOMP_MAP_ALWAYS_TOFROM: + OMP_CLAUSE_SET_MAP_KIND (*pc, GOMP_MAP_ALWAYS_TO); + map_seen = 3; + break; case GOMP_MAP_FIRSTPRIVATE_POINTER: case GOMP_MAP_ALWAYS_POINTER: case GOMP_MAP_ATTACH_DETACH: @@ -21080,7 +21088,7 @@ c_parser_omp_target_enter_data (location_t loc, c_parser *parser, map_seen |= 1; error_at (OMP_CLAUSE_LOCATION (*pc), "%<#pragma omp target enter data%> with map-type other " - "than % or % on % clause"); + "than %, % or % on % clause"); *pc = OMP_CLAUSE_CHAIN (*pc); continue; } @@ -21159,6 +21167,14 @@ c_parser_omp_target_exit_data (location_t loc, c_parser *parser, case GOMP_MAP_DELETE: map_seen = 3; break; + case GOMP_MAP_TOFROM: + OMP_CLAUSE_SET_MAP_KIND (*pc, GOMP_MAP_FROM); + map_seen = 3; + break; + case GOMP_MAP_ALWAYS_TOFROM: + OMP_CLAUSE_SET_MAP_KIND (*pc, GOMP_MAP_ALWAYS_FROM); + map_seen = 3; + break; case GOMP_MAP_FIRSTPRIVATE_POINTER: case GOMP_MAP_ALWAYS_POINTER: case GOMP_MAP_ATTACH_DETACH: @@ -21167,8 +21183,8 @@ c_parser_omp_target_exit_data (location_t loc, c_parser *parser, map_seen |= 1; error_at (OMP_CLAUSE_LOCATION (*pc), "%<#pragma omp target exit data%> with map-type other " - "than %, % or % on %" - " clause"); + "than %, %, % or % " + "on % clause"); *pc = OMP_CLAUSE_CHAIN (*pc); continue; } -- cgit v1.1 From ed974488991256c50f151ccfb271e198072dfc4d Mon Sep 17 00:00:00 2001 From: GCC Administrator Date: Sat, 2 Jul 2022 00:16:26 +0000 Subject: Daily bump. --- gcc/c/ChangeLog | 6 ++++++ 1 file changed, 6 insertions(+) (limited to 'gcc/c') diff --git a/gcc/c/ChangeLog b/gcc/c/ChangeLog index 077b726..8a0b71a 100644 --- a/gcc/c/ChangeLog +++ b/gcc/c/ChangeLog @@ -1,3 +1,9 @@ +2022-07-01 Tobias Burnus + + * c-parser.cc (c_parser_omp_target_enter_data, + c_parser_omp_target_exit_data): Accept tofrom + map-type modifier but use 'to' / 'from' internally. + 2022-06-16 David Malcolm * c-decl.cc (implicitly_declare): Add auto_diagnostic_group to -- cgit v1.1 From 683f11843974f0bdf42f79cdcbb0c2b43c7b81b0 Mon Sep 17 00:00:00 2001 From: Tobias Burnus Date: Mon, 4 Jul 2022 13:51:02 +0200 Subject: OpenMP: Move omp requires checks to libgomp Handle reverse_offload, unified_address, and unified_shared_memory requirements in libgomp by saving them alongside the offload table. When the device lto1 runs, it extracts the data for mkoffload. The latter than passes the value on to GOMP_offload_register_ver. lto1 (either the host one, with -flto [+ ENABLE_OFFLOADING], or in the offload-device lto1) also does the the consistency check is done, erroring out when the 'omp requires' clause use is inconsistent. For all in-principle supported devices, if a requirement cannot be fulfilled, the device is excluded from the (supported) devices list. Currently, none of those requirements are marked as supported for any of the non-host devices. gcc/c/ChangeLog: * c-parser.cc (c_parser_omp_target_data, c_parser_omp_target_update, c_parser_omp_target_enter_data, c_parser_omp_target_exit_data): Set OMP_REQUIRES_TARGET_USED. (c_parser_omp_requires): Remove sorry. gcc/ChangeLog: * config/gcn/mkoffload.cc (process_asm): Write '#include '. (process_obj): Pass omp_requires_mask to GOMP_offload_register_ver. (main): Ask lto1 to obtain omp_requires_mask and pass it on. * config/nvptx/mkoffload.cc (process, main): Likewise. * lto-cgraph.cc (omp_requires_to_name): New. (input_offload_tables): Save omp_requires_mask. (output_offload_tables): Read it, check for consistency, save value for mkoffload. * omp-low.cc (lower_omp_target): Force output_offloadtables call for OMP_REQUIRES_TARGET_USED. gcc/cp/ChangeLog: * parser.cc (cp_parser_omp_target_data, cp_parser_omp_target_enter_data, cp_parser_omp_target_exit_data, cp_parser_omp_target_update): Set OMP_REQUIRES_TARGET_USED. (cp_parser_omp_requires): Remove sorry. gcc/fortran/ChangeLog: * openmp.cc (gfc_match_omp_requires): Remove sorry. * parse.cc (decode_omp_directive): Don't regard 'declare target' as target usage for 'omp requires'; add more flags to omp_requires_mask. include/ChangeLog: * gomp-constants.h (GOMP_VERSION): Bump to 2. (GOMP_REQUIRES_UNIFIED_ADDRESS, GOMP_REQUIRES_UNIFIED_SHARED_MEMORY, GOMP_REQUIRES_REVERSE_OFFLOAD, GOMP_REQUIRES_TARGET_USED): New defines. libgomp/ChangeLog: * libgomp-plugin.h (GOMP_OFFLOAD_get_num_devices): Add omp_requires_mask arg. * plugin/plugin-gcn.c (GOMP_OFFLOAD_get_num_devices): Likewise; return -1 when device available but omp_requires_mask != 0. * plugin/plugin-nvptx.c (GOMP_OFFLOAD_get_num_devices): Likewise. * oacc-host.c (host_get_num_devices, host_openacc_get_property): Update call. * oacc-init.c (resolve_device, acc_init_1, acc_shutdown_1, goacc_attach_host_thread_to_device, acc_get_num_devices, acc_set_device_num, get_property_any): Likewise. * target.c (omp_requires_mask): New global var. (gomp_requires_to_name): New. (GOMP_offload_register_ver): Handle passed omp_requires_mask. (gomp_target_init): Handle omp_requires_mask. * libgomp.texi (OpenMP 5.0): Update requires impl. status. (OpenMP 5.1): Add a missed item. (OpenMP 5.2): Mark linear-clause change as supported in C/C++. * testsuite/libgomp.c-c++-common/requires-1-aux.c: New test. * testsuite/libgomp.c-c++-common/requires-1.c: New test. * testsuite/libgomp.c-c++-common/requires-2-aux.c: New test. * testsuite/libgomp.c-c++-common/requires-2.c: New test. * testsuite/libgomp.c-c++-common/requires-3-aux.c: New test. * testsuite/libgomp.c-c++-common/requires-3.c: New test. * testsuite/libgomp.c-c++-common/requires-4-aux.c: New test. * testsuite/libgomp.c-c++-common/requires-4.c: New test. * testsuite/libgomp.c-c++-common/requires-5-aux.c: New test. * testsuite/libgomp.c-c++-common/requires-5.c: New test. * testsuite/libgomp.c-c++-common/requires-6.c: New test. * testsuite/libgomp.c-c++-common/requires-7-aux.c: New test. * testsuite/libgomp.c-c++-common/requires-7.c: New test. * testsuite/libgomp.fortran/requires-1-aux.f90: New test. * testsuite/libgomp.fortran/requires-1.f90: New test. liboffloadmic/ChangeLog: * plugin/libgomp-plugin-intelmic.cpp (GOMP_OFFLOAD_get_num_devices): Return -1 when device available but omp_requires_mask != 0. gcc/testsuite/ChangeLog: * c-c++-common/gomp/requires-4.c: Update dg-*. * c-c++-common/gomp/reverse-offload-1.c: Likewise. * c-c++-common/gomp/target-device-ancestor-2.c: Likewise. * c-c++-common/gomp/target-device-ancestor-3.c: Likewise. * c-c++-common/gomp/target-device-ancestor-4.c: Likewise. * c-c++-common/gomp/target-device-ancestor-5.c: Likewise. * gfortran.dg/gomp/target-device-ancestor-3.f90: Likewise. * gfortran.dg/gomp/target-device-ancestor-4.f90: Likewise. * gfortran.dg/gomp/target-device-ancestor-5.f90: Likewise. * gfortran.dg/gomp/target-device-ancestor-2.f90: Likewise. Move post-FE checks to ... * gfortran.dg/gomp/target-device-ancestor-2a.f90: ... this new file. * gfortran.dg/gomp/requires-8.f90: Update as we don't regard 'declare target' for the 'requires' usage requirement. Co-authored-by: Chung-Lin Tang Co-authored-by: Thomas Schwinge --- gcc/c/c-parser.cc | 19 ++++++++++++++++--- 1 file changed, 16 insertions(+), 3 deletions(-) (limited to 'gcc/c') diff --git a/gcc/c/c-parser.cc b/gcc/c/c-parser.cc index 97e3b23..9c02141 100644 --- a/gcc/c/c-parser.cc +++ b/gcc/c/c-parser.cc @@ -20915,6 +20915,10 @@ c_parser_omp_teams (location_t loc, c_parser *parser, static tree c_parser_omp_target_data (location_t loc, c_parser *parser, bool *if_p) { + if (flag_openmp) + omp_requires_mask + = (enum omp_requires) (omp_requires_mask | OMP_REQUIRES_TARGET_USED); + tree clauses = c_parser_omp_all_clauses (parser, OMP_TARGET_DATA_CLAUSE_MASK, "#pragma omp target data"); @@ -21010,6 +21014,10 @@ c_parser_omp_target_update (location_t loc, c_parser *parser, return false; } + if (flag_openmp) + omp_requires_mask + = (enum omp_requires) (omp_requires_mask | OMP_REQUIRES_TARGET_USED); + tree stmt = make_node (OMP_TARGET_UPDATE); TREE_TYPE (stmt) = void_type_node; OMP_TARGET_UPDATE_CLAUSES (stmt) = clauses; @@ -21057,6 +21065,10 @@ c_parser_omp_target_enter_data (location_t loc, c_parser *parser, return true; } + if (flag_openmp) + omp_requires_mask + = (enum omp_requires) (omp_requires_mask | OMP_REQUIRES_TARGET_USED); + tree clauses = c_parser_omp_all_clauses (parser, OMP_TARGET_ENTER_DATA_CLAUSE_MASK, "#pragma omp target enter data"); @@ -21151,6 +21163,10 @@ c_parser_omp_target_exit_data (location_t loc, c_parser *parser, return true; } + if (flag_openmp) + omp_requires_mask + = (enum omp_requires) (omp_requires_mask | OMP_REQUIRES_TARGET_USED); + tree clauses = c_parser_omp_all_clauses (parser, OMP_TARGET_EXIT_DATA_CLAUSE_MASK, "#pragma omp target exit data"); @@ -22779,9 +22795,6 @@ c_parser_omp_requires (c_parser *parser) c_parser_skip_to_pragma_eol (parser, false); return; } - if (p && this_req != OMP_REQUIRES_DYNAMIC_ALLOCATORS) - sorry_at (cloc, "%qs clause on % directive not " - "supported yet", p); if (p) c_parser_consume_token (parser); if (this_req) -- cgit v1.1 From 8467574d8daac47e0cf5b694f6c012aad8d630a6 Mon Sep 17 00:00:00 2001 From: GCC Administrator Date: Tue, 5 Jul 2022 00:16:36 +0000 Subject: Daily bump. --- gcc/c/ChangeLog | 9 +++++++++ 1 file changed, 9 insertions(+) (limited to 'gcc/c') diff --git a/gcc/c/ChangeLog b/gcc/c/ChangeLog index 8a0b71a..989f293 100644 --- a/gcc/c/ChangeLog +++ b/gcc/c/ChangeLog @@ -1,3 +1,12 @@ +2022-07-04 Tobias Burnus + Chung-Lin Tang + Thomas Schwinge + + * c-parser.cc (c_parser_omp_target_data, c_parser_omp_target_update, + c_parser_omp_target_enter_data, c_parser_omp_target_exit_data): Set + OMP_REQUIRES_TARGET_USED. + (c_parser_omp_requires): Remove sorry. + 2022-07-01 Tobias Burnus * c-parser.cc (c_parser_omp_target_enter_data, -- cgit v1.1