From: Tobias Burnus Date: Wed, 14 Jun 2023 05:53:02 +0000 (+0200) Subject: OpenMP: Set default-device-var with OMP_TARGET_OFFLOAD=mandatory X-Git-Tag: basepoints/gcc-15~8338 X-Git-Url: http://git.ipfire.org/cgi-bin/gitweb.cgi?a=commitdiff_plain;h=18c8b56c7d67a9e37acf28822587786f0fc0efbc;p=thirdparty%2Fgcc.git OpenMP: Set default-device-var with OMP_TARGET_OFFLOAD=mandatory OMP_TARGET_OFFLOAD=mandatory handling was before inconsistent. Hence, in OpenMP 5.2 it was clarified/extended by having implications on the default-device-var; additionally, omp_initial_device and omp_invalid_device enum values/PARAMETERs were added; support for it was added in r13-1066-g1158fe43407568 including aborting for omp_invalid_device and non-conforming device numbers. Only the mandatory handling was missing. Namely, while the default-device-var is usually initialized to value 0, with 'mandatory' it must have the value 'omp_invalid_device' if and only if zero non-host devices are available. (The OMP_DEFAULT_DEVICE env var overrides this as it comes semantically after the initialization.) To achieve this, default-device-var is now initialized to MIN_INT. If there is no 'mandatory', it is set to 0 directly after env var parsing. Otherwise, it is updated in gomp_target_init to either 0 or omp_invalid_device. To ensure INT_MIN is never seen by the user, both the omp_get_default_device API routine and omp_display_env (user call and OMP_DISPLAY_ENV env var) call gomp_init_targets_once() in that case. libgomp/ChangeLog: * env.c (gomp_default_icv_values): Init default_device_var to an nonconforming value - INT_MIN. (initialize_env): After env-var parsing, set default_device_var to device 0 unless OMP_TARGET_OFFLOAD=mandatory. (omp_display_env): If default_device_var is INT_MIN, call gomp_init_targets_once. * icv-device.c (omp_get_default_device): Likewise. * libgomp.texi (OMP_DEFAULT_DEVICE): Update init description. (OpenMP 5.2 Impl. Status): Mark OMP_TARGET_OFFLOAD=mandatory as 'Y'. * target.c (resolve_device): Improve error message device-num < 0 with 'mandatory' and no no-host devices available. (gomp_target_init): Set default-device-var if INT_MIN. * testsuite/libgomp.c/target-48.c: New test. * testsuite/libgomp.c/target-49.c: New test. * testsuite/libgomp.c/target-50.c: New test. * testsuite/libgomp.c/target-50a.c: New test. * testsuite/libgomp.c/target-51.c: New test. * testsuite/libgomp.c/target-52.c: New test. * testsuite/libgomp.c/target-53.c: New test. * testsuite/libgomp.c/target-54.c: New test. --- diff --git a/libgomp/env.c b/libgomp/env.c index e7a035b593c7..25c0211dda16 100644 --- a/libgomp/env.c +++ b/libgomp/env.c @@ -62,13 +62,14 @@ #include "secure_getenv.h" #include "environ.h" -/* Default values of ICVs according to the OpenMP standard. */ +/* Default values of ICVs according to the OpenMP standard, + except for default-device-var. */ const struct gomp_default_icv gomp_default_icv_values = { .nthreads_var = 1, .thread_limit_var = UINT_MAX, .run_sched_var = GFS_DYNAMIC, .run_sched_chunk_size = 1, - .default_device_var = 0, + .default_device_var = INT_MIN, .max_active_levels_var = 1, .bind_var = omp_proc_bind_false, .nteams_var = 0, @@ -1614,6 +1615,10 @@ omp_display_env (int verbose) struct gomp_icv_list *none = gomp_get_initial_icv_item (GOMP_DEVICE_NUM_FOR_NO_SUFFIX); + if (none->icvs.default_device_var == INT_MIN) + /* This implies OMP_TARGET_OFFLOAD=mandatory. */ + gomp_init_targets_once (); + fputs ("\nOPENMP DISPLAY ENVIRONMENT BEGIN\n", stderr); fputs (" _OPENMP = '201511'\n", stderr); @@ -2213,6 +2218,10 @@ initialize_env (void) gomp_global_icv.max_active_levels_var = gomp_supported_active_levels; } + if (gomp_global_icv.default_device_var == INT_MIN + && gomp_target_offload_var != GOMP_TARGET_OFFLOAD_MANDATORY) + none->icvs.default_device_var = gomp_global_icv.default_device_var = 0; + /* Process GOMP_* variables and dependencies between parsed ICVs. */ parse_int_secure ("GOMP_DEBUG", &gomp_debug_var, true); diff --git a/libgomp/icv-device.c b/libgomp/icv-device.c index a2bbedc672a4..b48ea3b096cf 100644 --- a/libgomp/icv-device.c +++ b/libgomp/icv-device.c @@ -27,6 +27,7 @@ expected to replace. */ #include "libgomp.h" +#include void omp_set_default_device (int device_num) @@ -41,6 +42,9 @@ int omp_get_default_device (void) { struct gomp_task_icv *icv = gomp_icv (false); + if (icv->default_device_var == INT_MIN) + /* This implies OMP_TARGET_OFFLOAD=mandatory. */ + gomp_init_targets_once (); return icv->default_device_var; } diff --git a/libgomp/libgomp.texi b/libgomp/libgomp.texi index a3d370a0fb33..21d3582a6657 100644 --- a/libgomp/libgomp.texi +++ b/libgomp/libgomp.texi @@ -423,7 +423,7 @@ to address of matching mapped list item per 5.1, Sect. 2.21.7.2 @tab N @tab @item Conforming device numbers and @code{omp_initial_device} and @code{omp_invalid_device} enum/PARAMETER @tab Y @tab @item Initial value of @emph{default-device-var} ICV with - @code{OMP_TARGET_OFFLOAD=mandatory} @tab N @tab + @code{OMP_TARGET_OFFLOAD=mandatory} @tab Y @tab @item @emph{interop_types} in any position of the modifier list for the @code{init} clause of the @code{interop} construct @tab N @tab @end multitable @@ -2006,6 +2006,8 @@ Set to choose the device which is used in a @code{target} region, unless the value is overridden by @code{omp_set_default_device} or by a @code{device} clause. The value shall be the nonnegative device number. If no device with the given device number exists, the code is executed on the host. If unset, +@env{OMP_TARGET_OFFLOAD} is @code{mandatory} and no non-host devices are +available, it is set to @code{omp_invalid_device}. Otherwise, if unset, device number 0 will be used. diff --git a/libgomp/target.c b/libgomp/target.c index e3c4121a09ff..f1020fad601b 100644 --- a/libgomp/target.c +++ b/libgomp/target.c @@ -150,7 +150,11 @@ resolve_device (int device_id, bool remapped) if (device_id == (remapped ? GOMP_DEVICE_HOST_FALLBACK : omp_initial_device)) return NULL; - if (device_id == omp_invalid_device) + if (gomp_target_offload_var == GOMP_TARGET_OFFLOAD_MANDATORY + && gomp_get_num_devices () == 0) + gomp_fatal ("OMP_TARGET_OFFLOAD is set to MANDATORY but only the host " + "device is available"); + else if (device_id == omp_invalid_device) gomp_fatal ("omp_invalid_device encountered"); else if (gomp_target_offload_var == GOMP_TARGET_OFFLOAD_MANDATORY) gomp_fatal ("OMP_TARGET_OFFLOAD is set to MANDATORY, " @@ -5184,6 +5188,15 @@ gomp_target_init (void) if (devs[i].capabilities & GOMP_OFFLOAD_CAP_OPENACC_200) goacc_register (&devs[i]); } + if (gomp_global_icv.default_device_var == INT_MIN) + { + /* This implies OMP_TARGET_OFFLOAD=mandatory. */ + struct gomp_icv_list *none; + none = gomp_get_initial_icv_item (GOMP_DEVICE_NUM_FOR_NO_SUFFIX); + gomp_global_icv.default_device_var = (num_devs_openmp + ? 0 : omp_invalid_device); + none->icvs.default_device_var = gomp_global_icv.default_device_var; + } num_devices = num_devs; num_devices_openmp = num_devs_openmp; diff --git a/libgomp/testsuite/libgomp.c/target-48.c b/libgomp/testsuite/libgomp.c/target-48.c new file mode 100644 index 000000000000..8e95c1c3ac38 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-48.c @@ -0,0 +1,31 @@ +/* Check OMP_TARGET_OFFLOAD on systems with no available non-host devices; + omp_invalid_device == -4 with GCC. */ + +/* { dg-do run { target { ! offload_device } } } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */ + +/* { dg-output ".*OMP_DEFAULT_DEVICE = '-4'.*OMP_TARGET_OFFLOAD = 'MANDATORY'.*" } */ + +#include + +int +main () +{ + if (omp_get_default_device () != omp_invalid_device) + __builtin_abort (); + + omp_set_default_device (omp_initial_device); + + /* The spec is a bit unclear whether the line above sets the device number + (a) to -1 (= omp_initial_device) or + (b) to omp_get_initial_device() == omp_get_num_devices(). Therefore, + we accept either value. */ + + if (omp_get_default_device() != omp_get_initial_device() + && omp_get_default_device() != omp_initial_device) + __builtin_abort (); + + omp_display_env (0); + + return 0; +} diff --git a/libgomp/testsuite/libgomp.c/target-49.c b/libgomp/testsuite/libgomp.c/target-49.c new file mode 100644 index 000000000000..970cb91d5124 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-49.c @@ -0,0 +1,18 @@ +/* Check OMP_TARGET_OFFLOAD on systems with no available non-host devices, + which is enforced by using -foffload=disable. */ + +/* { dg-do run } */ +/* { dg-additional-options "-foffload=disable" } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */ +/* { dg-set-target-env-var OMP_DISPLAY_ENV "true" } */ + +/* See comment in target-50.c/target-50.c for why default-device-var can be '0'. */ + +/* { dg-output ".*OMP_DEFAULT_DEVICE = '-4'.*OMP_TARGET_OFFLOAD = 'MANDATORY'.*" { target { ! offload_device } } } */ +/* { dg-output ".*OMP_DEFAULT_DEVICE = '0'.*OMP_TARGET_OFFLOAD = 'MANDATORY'.*" { target offload_device } } */ + +int +main () +{ + return 0; +} diff --git a/libgomp/testsuite/libgomp.c/target-50.c b/libgomp/testsuite/libgomp.c/target-50.c new file mode 100644 index 000000000000..6f15569ee211 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-50.c @@ -0,0 +1,27 @@ +/* Check OMP_TARGET_OFFLOAD on systems with no available non-host devices; + here with using -foffload=disable. + As default-device-var is set to 0 (= host in this case), it should not fail. */ + +/* Note that -foffload=disable will still find devices on the system and only + when trying to use them, it will fail as no binary data has been produced. + The "target offload_device" case is checked for in 'target-50a.c'. */ + +/* { dg-do run { target { ! offload_device } } } */ + +/* { dg-additional-options "-foffload=disable" } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */ +/* { dg-set-target-env-var OMP_DEFAULT_DEVICE "0" } */ +/* { dg-set-target-env-var OMP_DISPLAY_ENV "true" } */ + +/* { dg-output ".*OMP_DEFAULT_DEVICE = '0'.*OMP_TARGET_OFFLOAD = 'MANDATORY'.*" } */ + +int +main () +{ + int x; + #pragma omp target map(tofrom:x) + x = 5; + if (x != 5) + __builtin_abort (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c/target-50a.c b/libgomp/testsuite/libgomp.c/target-50a.c new file mode 100644 index 000000000000..0835cb5bae35 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-50a.c @@ -0,0 +1,43 @@ +/* Check OMP_TARGET_OFFLOAD on systems with non-host devices but no executable + code due to -foffload=disable. + + Note: While one might expect that -foffload=disable implies no non-host + devices, libgomp actually detects the devices and only fails when trying to + run as no executable code is availale for that device. + (Without MANDATORY it simply uses host fallback, which should usually be fine + but might have issues in corner cases.) + + We have default-device-var = 0 (default but also explicitly set), which will + fail at runtime. For -foffload=disable without non-host devices, see + target-50.c testcase. */ + +/* { dg-do run { target offload_device } } */ + +/* { dg-additional-options "-foffload=disable" } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */ +/* { dg-set-target-env-var OMP_DEFAULT_DEVICE "0" } */ +/* { dg-set-target-env-var OMP_DISPLAY_ENV "true" } */ + +/* { dg-output ".*OMP_DEFAULT_DEVICE = '0'.*OMP_TARGET_OFFLOAD = 'MANDATORY'.*" } */ + +#include + +int +main () +{ + int x; + /* We know that there are non-host devices. With GCC, we still find them as + available devices, hence, check for it. */ + if (omp_get_num_devices() <= 0) + __builtin_abort (); + + /* But due to -foffload=disable, there are no binary code for (default) device '0' */ + + /* { dg-output ".*libgomp: OMP_TARGET_OFFLOAD is set to MANDATORY, but device cannot be used for offloading.*" } */ + /* { dg-shouldfail "OMP_TARGET_OFFLOAD=mandatory and no binary code for a non-host device" } */ + #pragma omp target map(tofrom:x) + x = 5; + if (x != 5) + __builtin_abort (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c/target-51.c b/libgomp/testsuite/libgomp.c/target-51.c new file mode 100644 index 000000000000..7d09bceacd58 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-51.c @@ -0,0 +1,24 @@ +/* Check OMP_TARGET_OFFLOAD on systems with no available non-host devices, + which is enforced by using -foffload=disable. */ + +/* { dg-do run } */ +/* { dg-additional-options "-foffload=disable" } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */ + +/* { dg-shouldfail "OMP_TARGET_OFFLOAD=mandatory and no available device" } */ + +/* See comment in target-50.c/target-50.c for why the output differs. */ + +/* { dg-output ".*libgomp: OMP_TARGET_OFFLOAD is set to MANDATORY but only the host device is available.*" { target { ! offload_device } } } */ +/* { dg-output ".*libgomp: OMP_TARGET_OFFLOAD is set to MANDATORY but device not found.*" { target offload_device } } */ + +int +main () +{ + int x; + #pragma omp target map(tofrom:x) + x = 5; + if (x != 5) + __builtin_abort (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c/target-52.c b/libgomp/testsuite/libgomp.c/target-52.c new file mode 100644 index 000000000000..809380c6928d --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-52.c @@ -0,0 +1,25 @@ +/* Only run this with available non-host devices; in that case, GCC sets + the default-device-var to 0. */ + +/* { dg-do run { target { offload_device } } } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */ +/* { dg-set-target-env-var OMP_DISPLAY_ENV "true" } */ + +/* { dg-output ".*OMP_DEFAULT_DEVICE = '0'.*OMP_TARGET_OFFLOAD = 'MANDATORY'.*" } */ + +#include + +int +main () +{ + int x; + #pragma omp target map(tofrom:x) + x = 5 + omp_is_initial_device (); + + if (x != 5) + __builtin_abort (); + + if (0 != omp_get_default_device()) + __builtin_abort (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c/target-53.c b/libgomp/testsuite/libgomp.c/target-53.c new file mode 100644 index 000000000000..866e8961af19 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-53.c @@ -0,0 +1,22 @@ +/* { dg-do run } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "disabled" } */ +/* { dg-set-target-env-var OMP_DISPLAY_ENV "true" } */ + +/* { dg-output ".*OMP_DEFAULT_DEVICE = '\[0-9\]+'.*OMP_TARGET_OFFLOAD = 'DISABLED'.*" } */ + +#include + +int +main () +{ + int x; + #pragma omp target map(tofrom:x) + x = 5 + omp_is_initial_device (); + + if (x != 5+1) + __builtin_abort (); + + if (omp_get_default_device() != omp_get_initial_device()) + __builtin_abort (); + return 0; +} diff --git a/libgomp/testsuite/libgomp.c/target-54.c b/libgomp/testsuite/libgomp.c/target-54.c new file mode 100644 index 000000000000..bc4e69b52787 --- /dev/null +++ b/libgomp/testsuite/libgomp.c/target-54.c @@ -0,0 +1,20 @@ +/* { dg-do run } */ +/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "default" } */ +/* { dg-set-target-env-var OMP_DISPLAY_ENV "true" } */ + +/* { dg-output ".*OMP_DEFAULT_DEVICE = '0'.*OMP_TARGET_OFFLOAD = 'DEFAULT'.*" } */ + +#include + +int +main () +{ + int x; + #pragma omp target map(tofrom:x) + x = 5 + omp_is_initial_device (); + + if (x != 5 + (omp_get_default_device() == omp_get_initial_device())) + __builtin_abort (); + + return 0; +}