From: Marcel Vollweiler <marcel@codesourcery.com>
To: Jakub Jelinek <jakub@redhat.com>
Cc: <gcc-patches@gcc.gnu.org>
Subject: Re: [PATCH] OpenMP, libgomp: Environment variable syntax extension.
Date: Tue, 2 Aug 2022 09:52:02 +0200 [thread overview]
Message-ID: <c33653d4-20e7-96c3-4b9f-61edf2628682@codesourcery.com> (raw)
In-Reply-To: <055f7cfb-2849-ba5a-a161-13333e19e538@codesourcery.com>
[-- Attachment #1: Type: text/plain, Size: 23997 bytes --]
Hi Jakub,
I updated (and simplified) my last submitted patch
(https://gcc.gnu.org/pipermail/gcc-patches/2022-July/598770.html) considering
the following aspects:
1. For copying ICVs to devices a struct ("gomp_offload_icvs") is used now
instead of copying all ICVs individually. This is a somehow reduced struct as we
don't want to copy ICVs which are not used on the device.
2. A linked list of those gomp_offload_icvs structs was introduced
(gomp_offload_icv_list) that holds only device-specific ICV values. In a
following patch (for
https://gcc.gnu.org/pipermail/gcc-patches/2022-April/593260.html, will be
submitted soon) this list is also used to keep ICV values that were changed on a
particular device and copied back to the host.
3. The value of a device-specific ICV according to the hierarchy (DEV_X, DEV,
ALL) can be determined earliest when we copy to the device, since we do not know
the actual existing device numbers (only those that were explicitly specified).
The device-specific ICV struct is added to gomp_offload_icv_list when it is
copied the first time. Once it is contained it is taken from the list for each
copy process (to the device and back).
The updated patch is attached and tested again on x86_64-linux with nvptx and
amdgcn offloading without regression.
Marcel
Am 25.07.2022 um 15:38 schrieb Marcel Vollweiler:
> Hi Jakub,
>
>>>> I'm not sure we can rely on execv on all targets that do support libgomp.
>>>> Any reason why you actually need this, rather than using
>>>> dg-set-target-env-var directive(s) and perhaps return 0; if getenv doesn't
>>>> return the expected values?
>>>
>>> Interesting topic. After some (internal) discussions I think the best way is to
>>> set the environment variables explicitely instead using dg-set-target-env-var.
>>> The reason is that dg-set-target-env-var does not work for remote testing (which
>>> seems to be a common test environment). For remote testing dejagnu immediately
>>> aborts the test case with UNSUPPORTED which is specified in the corresponding
>>> extension and makes sence from my point of view as the test assumption cannot be
>>> fulfilled (since the environment variables are not set on remote targets).
>>> It also means that whenever dg-set-target-env-var is set in the test file, the
>>> execution of the test case is not tested on remote targets.
>>
>> The only reason why dg-set-target-env-var is supported on native only right
>> now is that I'm never doing remote testing myself and so couldn't test that.
>> There is no inherent reason why the env vars couldn't be propagated over to
>> the remote and set in the environment there.
>> So trying to work around that rather than at least trying to change
>> dg-set-target-env-var so that it works with the remote testing you do looks
>> wrong.
>> If dg-set-target-env-var can be made to work remotely, it will magically
>> improve those 130+ tests that use it already together with the newly added
>> tests.
>>
>> So, I'd suggest to just use dg-set-target-env-var and incrementally work on
>> making it work for remote testing if that is important to whomever does
>> that kind of testing. Could be e.g. a matter of invoking remotely
>> env VAR1=val1 VAR2=val2 program args
>> instead of program args. If env is missing on the remote side, it could
>> be UNSUPPORTED then.
>
> I agree. So I changed the tests using dg-set-target-env-var and removed the
> execv parts.
>
>>
>>> +/* The initial ICV values for the host, which are configured with environment
>>> + variables without a suffix, e.g. OMP_NUM_TEAMS. */
>>> +struct gomp_initial_icvs gomp_initial_icvs_none;
>>> +
>>> +/* Initial ICV values that were configured for the host and for all devices by
>>> + using environment variables like OMP_NUM_TEAMS_ALL. */
>>> +struct gomp_initial_icvs gomp_initial_icvs_all;
>>> +
>>> +/* Initial ICV values that were configured only for devices (not for the host)
>>> + by using environment variables like OMP_NUM_TEAMS_DEV. */
>>> +struct gomp_initial_icvs gomp_initial_icvs_dev;
>>
>> As I said last time, I don't like allocating these
>> all the time in the data section of libgomp when at least for a few upcoming
>> years, most users will never use those suffixes.
>> Can't *_DEV and *_ALL go into the gomp_initial_icv_dev_list
>> chain too, perhaps
>
> gomp_initial_icvs_{none, all, dev} are now defined as pointers (as you proposed
> previously). gomp_initial_icvs_{all, dev} are only instantiated if at least one
> according environment variable is parsed. gomp_initial_icvs_none is always
> initialized with the initial global ICV values.
>
> All three structures are now also included in gomp_initial_icv_list (previously
> named gomp_initial_icv_dev_list) with "magic device numbers" -1, -2, and -3.
> The list items for _DEV, _ALL and no suffix are stored at the beginning of the
> list whereas the device-specific list items are attached at the end.
>
>>
>>> +static const struct envvar
>>> +{
>>> + const char *name;
>>> + int name_len;
>>> + unsigned char flag_vars[3];
>>> + unsigned char flag;
>>> + void *params[3];
>>> + bool (*parse_func) (const char *, const char *, void * const[]);
>>> +} envvars[] = {
>>> + {ENTRY ("OMP_SCHEDULE_DEV"), {OMP_SCHEDULE_DEV_,
>>> OMP_SCHEDULE_CHUNK_SIZE_DEV_}, GOMP_ENV_VAR_SUFFIX_DEV,
>>> {&gomp_initial_icvs_dev.run_sched_var,
>>> &gomp_initial_icvs_dev.run_sched_chunk_size}, &parse_schedule},
>>> + {ENTRY ("OMP_SCHEDULE_ALL"), {OMP_SCHEDULE_DEV_,
>>> OMP_SCHEDULE_CHUNK_SIZE_DEV_}, GOMP_ENV_VAR_SUFFIX_ALL,
>>> {&gomp_initial_icvs_all.run_sched_var,
>>> &gomp_initial_icvs_all.run_sched_chunk_size}, &parse_schedule},
>>> + {ENTRY ("OMP_SCHEDULE"), {OMP_SCHEDULE_DEV_,
>>> OMP_SCHEDULE_CHUNK_SIZE_DEV_}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_initial_icvs_none.run_sched_var,
>>> &gomp_initial_icvs_none.run_sched_chunk_size}, &parse_schedule},
>>> +
>>> + {ENTRY ("OMP_NUM_TEAMS_DEV"), {OMP_NUM_TEAMS_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_DEV , {&gomp_initial_icvs_dev.nteams_var, false},
>>> &parse_int},
>>> + {ENTRY ("OMP_NUM_TEAMS_ALL"), {OMP_NUM_TEAMS_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_ALL, {&gomp_initial_icvs_all.nteams_var, false},
>>> &parse_int},
>>> + {ENTRY ("OMP_NUM_TEAMS"), {OMP_NUM_TEAMS_DEV_}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_initial_icvs_none.nteams_var, false}, &parse_int},
>>> +
>>> + {ENTRY ("OMP_DYNAMIC_DEV"), {OMP_DYNAMIC_DEV_}, GOMP_ENV_VAR_SUFFIX_DEV,
>>> {&gomp_initial_icvs_dev.dyn_var}, &parse_boolean},
>>> + {ENTRY ("OMP_DYNAMIC_ALL"), {OMP_DYNAMIC_DEV_}, GOMP_ENV_VAR_SUFFIX_ALL,
>>> {&gomp_initial_icvs_all.dyn_var}, &parse_boolean},
>>> + {ENTRY ("OMP_DYNAMIC"), {OMP_DYNAMIC_DEV_}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_initial_icvs_none.dyn_var}, &parse_boolean},
>>> +
>>> + {ENTRY ("OMP_TEAMS_THREAD_LIMIT_DEV"), {OMP_TEAMS_THREAD_LIMIT_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_DEV, {&gomp_initial_icvs_dev.teams_thread_limit_var,
>>> false}, &parse_int},
>>> + {ENTRY ("OMP_TEAMS_THREAD_LIMIT_ALL"), {OMP_TEAMS_THREAD_LIMIT_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_ALL, {&gomp_initial_icvs_all.teams_thread_limit_var,
>>> false}, &parse_int},
>>> + {ENTRY ("OMP_TEAMS_THREAD_LIMIT"), {OMP_TEAMS_THREAD_LIMIT_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_NONE, {&gomp_initial_icvs_none.teams_thread_limit_var,
>>> false}, &parse_int},
>>> +
>>> + {ENTRY ("OMP_THREAD_LIMIT_DEV"), {OMP_THREAD_LIMIT_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_DEV, {&gomp_initial_icvs_dev.thread_limit_var, false,
>>> (void *) UINT_MAX}, &parse_unsigned_long},
>>> + {ENTRY ("OMP_THREAD_LIMIT_ALL"), {OMP_THREAD_LIMIT_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_ALL, {&gomp_initial_icvs_all.thread_limit_var, false,
>>> (void *) UINT_MAX}, &parse_unsigned_long},
>>> + {ENTRY ("OMP_THREAD_LIMIT"), {OMP_THREAD_LIMIT_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_NONE, {&gomp_initial_icvs_none.thread_limit_var, false,
>>> (void *) UINT_MAX}, &parse_unsigned_long},
>>> +
>>> + {ENTRY ("OMP_NUM_THREADS_DEV"), {OMP_NUM_THREADS_DEV_,
>>> OMP_NTHREADS_LIST_DEV, OMP_NTHREADS_LIST_LEN_DEV}, GOMP_ENV_VAR_SUFFIX_DEV,
>>> {&gomp_initial_icvs_dev.nthreads_var,
>>> &gomp_initial_icvs_dev.nthreads_var_list,
>>> &gomp_initial_icvs_dev.nthreads_var_list_len}, &parse_unsigned_long_list},
>>> + {ENTRY ("OMP_NUM_THREADS_ALL"), {OMP_NUM_THREADS_DEV_,
>>> OMP_NTHREADS_LIST_DEV, OMP_NTHREADS_LIST_LEN_DEV}, GOMP_ENV_VAR_SUFFIX_ALL,
>>> {&gomp_initial_icvs_all.nthreads_var,
>>> &gomp_initial_icvs_all.nthreads_var_list,
>>> &gomp_initial_icvs_all.nthreads_var_list_len}, &parse_unsigned_long_list},
>>> + {ENTRY ("OMP_NUM_THREADS"), {OMP_NUM_THREADS_DEV_, OMP_NTHREADS_LIST_DEV,
>>> OMP_NTHREADS_LIST_LEN_DEV}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_initial_icvs_none.nthreads_var,
>>> &gomp_initial_icvs_none.nthreads_var_list,
>>> &gomp_initial_icvs_none.nthreads_var_list_len}, &parse_unsigned_long_list},
>>> +
>>> + {ENTRY ("OMP_PROC_BIND_DEV"), {OMP_PROC_BIND_DEV_,
>>> OMP_PROC_BIND_LIST_DEV_, OMP_PROC_BIND_LIST_LEN_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_DEV, {&gomp_initial_icvs_dev.bind_var,
>>> &gomp_initial_icvs_dev.bind_var_list,
>>> &gomp_initial_icvs_dev.bind_var_list_len}, &parse_bind_var},
>>> + {ENTRY ("OMP_PROC_BIND_ALL"), {OMP_PROC_BIND_DEV_,
>>> OMP_PROC_BIND_LIST_DEV_, OMP_PROC_BIND_LIST_LEN_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_ALL, {&gomp_initial_icvs_all.bind_var,
>>> &gomp_initial_icvs_all.bind_var_list,
>>> &gomp_initial_icvs_all.bind_var_list_len}, &parse_bind_var},
>>> + {ENTRY ("OMP_PROC_BIND"), {OMP_PROC_BIND_DEV_, OMP_PROC_BIND_LIST_DEV_,
>>> OMP_PROC_BIND_LIST_LEN_DEV_}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_initial_icvs_none.bind_var, &gomp_initial_icvs_none.bind_var_list,
>>> &gomp_initial_icvs_none.bind_var_list_len}, &parse_bind_var},
>>> +
>>> + {ENTRY ("OMP_MAX_ACTIVE_LEVELS_DEV"), {OMP_MAX_ACTIVE_LEVELS_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_DEV, {&gomp_initial_icvs_dev.max_active_levels_var, (void
>>> *) true, (void *) gomp_supported_active_levels}, &parse_unsigned_long},
>>> + {ENTRY ("OMP_MAX_ACTIVE_LEVELS_ALL"), {OMP_MAX_ACTIVE_LEVELS_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_ALL, {&gomp_initial_icvs_all.max_active_levels_var, (void
>>> *) true, (void *) gomp_supported_active_levels}, &parse_unsigned_long},
>>> + {ENTRY ("OMP_MAX_ACTIVE_LEVELS"), {OMP_MAX_ACTIVE_LEVELS_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_NONE, {&gomp_initial_icvs_none.max_active_levels_var,
>>> (void *) true, (void *) gomp_supported_active_levels}, &parse_unsigned_long},
>>> +
>>> + {ENTRY ("OMP_WAIT_POLICY_DEV"), {OMP_WAIT_POLICY_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_DEV, {&gomp_initial_icvs_dev.wait_policy},
>>> &parse_wait_policy},
>>> + {ENTRY ("OMP_WAIT_POLICY_ALL"), {OMP_WAIT_POLICY_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_ALL, {&gomp_initial_icvs_all.wait_policy},
>>> &parse_wait_policy},
>>> + {ENTRY ("OMP_WAIT_POLICY"), {OMP_WAIT_POLICY_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_NONE, {&gomp_initial_icvs_none.wait_policy},
>>> &parse_wait_policy},
>>> +
>>> + {ENTRY ("OMP_STACKSIZE_DEV"), {OMP_STACKSIZE_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_DEV, {&gomp_initial_icvs_dev.stacksize}, &parse_stacksize},
>>> + {ENTRY ("OMP_STACKSIZE_ALL"), {OMP_STACKSIZE_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_ALL, {&gomp_initial_icvs_all.stacksize}, &parse_stacksize},
>>> + {ENTRY ("OMP_STACKSIZE"), {OMP_STACKSIZE_DEV_}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_initial_icvs_none.stacksize}, &parse_stacksize},
>>> +
>>> + {ENTRY ("OMP_CANCELLATION"), {}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_cancel_var}, &parse_boolean},
>>> + {ENTRY ("OMP_DISPLAY_AFFINITY"), {}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_display_affinity_var}, &parse_boolean},
>>> + {ENTRY ("OMP_TARGET_OFFLOAD"), {}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_target_offload_var}, &parse_target_offload},
>>> + {ENTRY ("OMP_MAX_TASK_PRIORITY"), {}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_max_task_priority_var, (void *) true}, &parse_int},
>>> + {ENTRY ("OMP_ALLOCATOR"), {}, GOMP_ENV_VAR_SUFFIX_NONE,
>>> {&gomp_def_allocator}, &parse_allocator},
>>> + {ENTRY ("OMP_DEFAULT_DEVICE"), {OMP_DEFAULT_DEVICE_DEV_},
>>> GOMP_ENV_VAR_SUFFIX_NONE, {&gomp_initial_icvs_none.default_device_var, (void
>>> *) true}, &parse_int},
>>> +};
>>
>> This is very large, why do you need 3 entries per most of the env vars?
>> Just one would be enough, with just some flag whether it accepts any
>> suffixes or not.
>
> The tables have been optimized now. Only one entry per ICV and removed some parts.
>
>> Lines are too long, they'd need wrapping.
>> I think the coding conventions are space after { and space before }.
>
> Both changed.
>
>>
>>> +static bool
>>> +get_device_num (char *env, int *dev_num, int *dev_num_len)
>>> +{
>>> + int pos = 0;
>>> + char dev_num_str[INT_MAX_STR_LEN+1];
>>> +
>>> + *dev_num_len = 0;
>>> + *dev_num = -1;
>>> + if (env == NULL)
>>> + return false;
>>> +
>>> + while (pos < INT_MAX_STR_LEN)
>>> + {
>>> + if (env[pos] == '\0' || env[pos] == '=')
>>> + break;
>>> + dev_num_str[pos] = env[pos];
>>> + pos++;
>>> + }
>>> +
>>> + if (env[pos] != '=' || (dev_num_str[0] == '0' && pos > 1))
>>> + return false;
>>> +
>>> + dev_num_str[pos] = '\0';
>>> + *dev_num = (int) strtoul (dev_num_str, 0, 10);
>>> + *dev_num_len = pos;
>>
>> Why do you copy the chars to a separate stack buffer?
>> strtoul will stop on anything that isn't a number.
>> So, just passing it second argument and verifying that
>> it points to '=' ('\0' wouldn't be valid) would be good enough
>> (or perhaps also verying it doesn't point to the start pointer
>> in case caller hasn't checked yet if it starts with a digit).
>
> Thanks for this hint. That makes it much simpler. Changed.
>
>> Also, the (int) cast means it throws away important information,
>> we'd treat say
>> 8589934593
>> on 64-bit arches as 1, I think we want to just ignore it.
>> Also, we should ignore negative values (though, for 5.2 there
>> is a question if OMP_NUM_THREADS_DEV_-1=5 is valid or not and
>> redundant with OMP_NUM_THREADS_DEV_4=5 (if omp_get_num_devices() == 5)
>> or OMP_NUM_THREADS and which one wins.
>
> Negative device numbers are rejected now. Btw. it seems that defining
> environment variables like "OMP_NUM_THREADS_DEV_-1" is not always valid due to
> "-" (invalid identifier name).
>
>>
>>> + return true;
>>> +}
>>> +
>>> +/* Helper function for initialize_env to add a device specific ICV value
>>> + to gomp_initial_icv_dev_list. */
>>> +
>>> +static void
>>> +add_device_specific_icv (int dev_num, int icv_code, void *value)
>>> +{
>>> + struct gomp_icv_list *list = gomp_initial_icv_dev_list;
>>> + while (list != NULL && list->device_num != dev_num)
>>> + list = list->next;
>>> +
>>> + if (list == NULL)
>>> + {
>>> + list =
>>> + (struct gomp_icv_list *) gomp_malloc (sizeof (struct gomp_icv_list));
>>
>> Formatting, = can't be at the end of line.
>>
>>> +static unsigned char
>>> +get_icv_flag (unsigned char flag_var)
>>> +{
>>> + switch (flag_var)
>>> + {
>>> + case OMP_NUM_TEAMS_DEV_:
>>> + return gomp_initial_icv_flags.nteams_var;
>>> + case OMP_SCHEDULE_DEV_:
>>> + return gomp_initial_icv_flags.run_sched_var;
>>> + case OMP_SCHEDULE_CHUNK_SIZE_DEV_:
>>> + return gomp_initial_icv_flags.run_sched_chunk_size;
>>> + case OMP_DYNAMIC_DEV_:
>>> + return gomp_initial_icv_flags.dyn_var;
>>> + case OMP_TEAMS_THREAD_LIMIT_DEV_:
>>> + return gomp_initial_icv_flags.teams_thread_limit_var;
>>> + case OMP_THREAD_LIMIT_DEV_:
>>> + return gomp_initial_icv_flags.thread_limit_var;
>>> + case OMP_NUM_THREADS_DEV_:
>>> + return gomp_initial_icv_flags.nthreads_var;
>>> + case OMP_NTHREADS_LIST_DEV:
>>> + return gomp_initial_icv_flags.nthreads_var_list;
>>> + case OMP_NTHREADS_LIST_LEN_DEV:
>>> + return gomp_initial_icv_flags.nthreads_var_list_len;
>>> + case OMP_PROC_BIND_DEV_:
>>> + return gomp_initial_icv_flags.bind_var;
>>> + case OMP_PROC_BIND_LIST_DEV_:
>>> + return gomp_initial_icv_flags.bind_var_list;
>>> + case OMP_PROC_BIND_LIST_LEN_DEV_:
>>> + return gomp_initial_icv_flags.bind_var_list_len;
>>> + case OMP_MAX_ACTIVE_LEVELS_DEV_:
>>> + return gomp_initial_icv_flags.max_active_levels_var;
>>> + case OMP_WAIT_POLICY_DEV_:
>>> + return gomp_initial_icv_flags.wait_policy;
>>> + case OMP_STACKSIZE_DEV_:
>>> + return gomp_initial_icv_flags.stacksize;
>>> + case OMP_DEFAULT_DEVICE_DEV_:
>>> + return gomp_initial_icv_flags.default_device_var;
>>> + default:
>>> + return OMP_NONE;
>>
>> Doesn't this function return a bitmask of GOMP_ENV_VAR_SUFFIX_WHATEVER
>> values? OMP_NONE isn't one of them, shouldn't it return 0 instead?
>> But more importantly, wouldn't it be easier if the icv_flags was just
>> an array indexed by flag_var? You don't need a large switch to handle
>> setting or getting it then. As you just want 4 bits per flag instead of 8,
>> you could index it by flag_var >> 1 and for flag_var & 1 shift right or left
>> by 4.
>
> The ICV flags are now defined as uint32_t. This is enough to store flags for our
> current 21 ICVs. The flags for _DEV, _DEV_X, _ALL and no suffix are now
> separated as we have the flags variable for each item of the initial icv list.
>
>>
>>> static void __attribute__((constructor))
>>> initialize_env (void)
>>> {
>>> - unsigned long thread_limit_var;
>>> - unsigned long max_active_levels_var;
>>> + extern char **environ;
>>> + char **env;
>>> + int omp_var, dev_num = 0, dev_num_len = 0, int_value, k;
>>> + bool bool_value, ignore = false;
>>> + char *env_val;
>>>
>>> /* Do a compile time check that mkomp_h.pl did good job. */
>>> omp_check_defines ();
>>>
>>> - parse_schedule ();
>>> - parse_boolean ("OMP_DYNAMIC", &gomp_global_icv.dyn_var);
>>> - parse_boolean ("OMP_CANCELLATION", &gomp_cancel_var);
>>> - parse_boolean ("OMP_DISPLAY_AFFINITY", &gomp_display_affinity_var);
>>> - parse_int ("OMP_DEFAULT_DEVICE", &gomp_global_icv.default_device_var, true);
>>> - parse_target_offload ("OMP_TARGET_OFFLOAD", &gomp_target_offload_var);
>>> - parse_int ("OMP_MAX_TASK_PRIORITY", &gomp_max_task_priority_var, true);
>>> - gomp_def_allocator = parse_allocator ();
>>> - if (parse_unsigned_long ("OMP_THREAD_LIMIT", &thread_limit_var, false))
>>> - {
>>> - gomp_global_icv.thread_limit_var
>>> - = thread_limit_var > INT_MAX ? UINT_MAX : thread_limit_var;
>>> - }
>>> - parse_int_secure ("GOMP_DEBUG", &gomp_debug_var, true);
>>> #ifndef HAVE_SYNC_BUILTINS
>>> gomp_mutex_init (&gomp_managed_threads_lock);
>>> #endif
>>> gomp_init_num_threads ();
>>> gomp_available_cpus = gomp_global_icv.nthreads_var;
>>> - if (!parse_unsigned_long_list ("OMP_NUM_THREADS",
>>> - &gomp_global_icv.nthreads_var,
>>> - &gomp_nthreads_var_list,
>>> - &gomp_nthreads_var_list_len))
>>> - gomp_global_icv.nthreads_var = gomp_available_cpus;
>>> - parse_int ("OMP_NUM_TEAMS", &gomp_nteams_var, false);
>>> - parse_int ("OMP_TEAMS_THREAD_LIMIT", &gomp_teams_thread_limit_var, false);
>>> - bool ignore = false;
>>> - if (parse_bind_var ("OMP_PROC_BIND",
>>> - &gomp_global_icv.bind_var,
>>> - &gomp_bind_var_list,
>>> - &gomp_bind_var_list_len)
>>> +
>>> + for (env = environ; *env != 0; env++)
>>> + {
>>> + if (!startswith (*env, "OMP_"))
>>> + continue;
>>
>> While the above is certainly good and quickly skips non-OpenMP env vars,
>> I think
>>
>>> +
>>> + for (omp_var = 0; omp_var < OMP_DEV_VAR_CNT; omp_var++)
>>> + {
>>> + if (startswith (*env, envvars_dev[omp_var].name))
>>
>> walking 36 entries for each OMP_ env var and doing strncmp for each is
>> expensive.
>>
>> Wouldn't it be better to just walk the name once, find out the suffix in
>> there and the length of the part before it too,
>> just a simple loop, stopping at '=' and when seeing '_', check if followed
>> by "ALL=", "DEV=" or "DEV_" followed by digit. That will determine the
>> kind and length of the env var name without the suffixes, so then
>> you can just walk the much shortened table with just 16 entries now
>> and can skip entries which don't have the computed length (the table
>> includes name_len, so start by
>> if (envvars[omp_var].name_len != name_len)
>> continue;
>> if (memcmp (*env + strlen ("OMP_"), envvars[omp_var].name,
>> envvars[omp_var].name_len) != 0)
>> continue;
>
> That's a good point. Together with the "envvars" table the parsing has been
> optimized. Similarly to you proposal I just use a flag in the table and parse
> the environment variable variants successively. First the basic variable name
> (e.g. OMP_NUM_TEAMS) and then checking whether "=", "_DEV=", "_ALL=", or "_DEV_"
> is following if allowed according to the table's flag. (Only) for a directly
> following "=" a check is not necessary because host variants are always allowed
> (thus GOMP_ENV_VAR_SUFFIX_NONE is also omitted in the table).
>
>>
>>> --- a/libgomp/libgomp.h
>>> +++ b/libgomp/libgomp.h
>>> @@ -454,6 +454,24 @@ struct gomp_team_state
>>>
>>> struct target_mem_desc;
>>>
>>> +#define OMP_NONE 0
>>> +#define OMP_NUM_TEAMS_DEV_ 1
>>> +#define OMP_SCHEDULE_DEV_ 2
>>> +#define OMP_SCHEDULE_CHUNK_SIZE_DEV_ 3
>>> +#define OMP_DYNAMIC_DEV_ 4
>>> +#define OMP_TEAMS_THREAD_LIMIT_DEV_ 5
>>> +#define OMP_THREAD_LIMIT_DEV_ 6
>>> +#define OMP_NUM_THREADS_DEV_ 7
>>> +#define OMP_NTHREADS_LIST_DEV 8
>>> +#define OMP_NTHREADS_LIST_LEN_DEV 9
>>> +#define OMP_PROC_BIND_DEV_ 10
>>> +#define OMP_PROC_BIND_LIST_DEV_ 11
>>> +#define OMP_PROC_BIND_LIST_LEN_DEV_ 12
>>> +#define OMP_MAX_ACTIVE_LEVELS_DEV_ 13
>>> +#define OMP_WAIT_POLICY_DEV_ 14
>>> +#define OMP_STACKSIZE_DEV_ 15
>>> +#define OMP_DEFAULT_DEVICE_DEV_ 16
>>
>> These aren't constans defined in OpenMP standard, so I think it would
>> be better to use different prefixes, say GOMP_ICV_WHATEVER or
>> GOMP_ENV_WHATEVER. OMP_NONE is very much non-descriptive of what
>> it means. Why do some defines have _DEV_ suffixes and others _DEV?
> They are renamed accordingly.
>
>> It should also be an enum rather than set of defines and I don't see a
>> reason for the _DEV* suffixes.
>
> Agreed and changed accordingly.
>
>>
>>> +
>>> /* These are the OpenMP 4.0 Internal Control Variables described in
>>> section 2.3.1. Those described as having one copy per task are
>>> stored within the structure; those described as having one copy
>>> @@ -473,6 +491,69 @@ struct gomp_task_icv
>>> struct target_mem_desc *target_data;
>>> };
>>>
>>> +#define GOMP_ENV_VAR_SUFFIX_UNKNOWN 0
>>> +#define GOMP_ENV_VAR_SUFFIX_NONE 1
>>> +#define GOMP_ENV_VAR_SUFFIX_DEV 2
>>> +#define GOMP_ENV_VAR_SUFFIX_ALL 4
>>> +#define GOMP_ENV_VAR_SUFFIX_DEV_X 8
>>
>> Similarly, make this an enum, and perhaps just GOMP_ENV_SUFFIX_WHATEVER ?
>
> Also changed.
>
> An updated patch is attached and tested again on x86_64-linux with nvptx and
> amdgcn offloading without regression.
>
> Marcel
> -----------------
> Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstraße 201, 80634
> München; Gesellschaft mit beschränkter Haftung; Geschäftsführer: Thomas Heurung,
> Frank Thürauf; Sitz der Gesellschaft: München; Registergericht München, HRB 106955
-----------------
Siemens Electronic Design Automation GmbH; Anschrift: Arnulfstraße 201, 80634 München; Gesellschaft mit beschränkter Haftung; Geschäftsführer: Thomas Heurung, Frank Thürauf; Sitz der Gesellschaft: München; Registergericht München, HRB 106955
[-- Attachment #2: extend-env-variables-syntax.diff --]
[-- Type: text/plain, Size: 93981 bytes --]
OpenMP, libgomp: Environment variable syntax extension.
This patch considers the environment variable syntax extension for
device-specific variants of environment variables from OpenMP 5.1 (see
OpenMP 5.1 specification, p. 75 and p. 639). An environment variable (e.g.
OMP_NUM_TEAMS) can have different suffixes:
_DEV (e.g. OMP_NUM_TEAMS_DEV): affects all devices but not the host.
_DEV_<device> (e.g. OMP_NUM_TEAMS_DEV_42): affects only device with
number <device>.
no suffix (e.g. OMP_NUM_TEAMS): affects only the host.
In future OpenMP versions also suffix _ALL will be introduced (see discussion
https://github.com/OpenMP/spec/issues/3179). This is also considered in this
patch:
_ALL (e.g. OMP_NUM_TEAMS_ALL): affects all devices and the host.
The precedence is as follows (descending). For the host:
1. no suffix
2. _ALL
For devices:
1. _DEV_<device>
2. _DEV
3. _ALL
That means, _DEV_<device> is used whenever available. Otherwise _DEV is used if
available, and at last _ALL. If there is no value for any of the variable
variants, default values are used as already implemented before.
This patch concerns parsing (a), storing (b), output (c) and transmission to the
device (d):
(a) The actual number of devices and the numbering are not known when parsing
the environment variables. Thus all environment variables are iterated and
searched for device-specific ones.
(b) Only configured device-specific variables are stored. Thus, linked lists
are used.
(c) The output is done in omp_display_env (see specification p. 468f). Global
ICVs are tagged with [all], see https://github.com/OpenMP/spec/issues/3179.
ICVs which are not global but aren't handled device-specific yet are tagged
with [host]. omp_display_env outputs the initial values of the ICVs. That's why
separate data structures are introduced (like gomp_initial_icv...).
(d) Device-specific ICVs which are already user accessible on the device are
transmitted to the device (moreover nteams-var is added and used for the tests).
libgomp/ChangeLog:
* config/gcn/icv-device.c (omp_get_default_device): Return device-
specific ICV.
(omp_get_max_teams): Added for GCN devices.
(omp_set_num_teams): Likewise.
(ialias): Likewise.
* config/nvptx/icv-device.c (omp_get_default_device): Return device-
specific ICV.
(omp_get_max_teams): Added for NVPTX devices.
(omp_set_num_teams): Likewise.
(ialias): Likewise.
* env.c (struct gomp_icv_list): New struct to store entries of initial
ICV values.
(struct gomp_offload_icv_list): New struct to store entries of device-
specific ICV values that are copied to the device and back.
(struct gomp_default_icv_t): New struct to store default values of ICVs
according to the OpenMP standard.
(parse_schedule): Generalized for different variants of OMP_SCHEDULE.
(print_env_var_error): Function that prints an error for invalid values
for ICVs.
(parse_unsigned_long_1): Removed getenv. Generalized.
(parse_unsigned_long): Likewise.
(parse_int_1): Likewise.
(parse_int): Likewise.
(parse_int_secure): Likewise.
(parse_unsigned_long_list): Likewise.
(parse_target_offload): Likewise.
(parse_bind_var): Likewise.
(parse_stacksize): Likewise.
(parse_boolean): Likewise.
(parse_wait_policy): Likewise.
(parse_allocator): Likewise.
(omp_display_env): Extended to output different variants of environment
variables.
(print_schedule): New helper function for omp_display_env which prints
the values of run_sched_var.
(print_proc_bind): New helper function for omp_display_env which prints
the values of proc_bind_var.
(enum gomp_parse_type): Collection of types used for parsing environment
variables.
(ENTRY): Preprocess string lengths of environment variables.
(OMP_VAR_CNT): Preprocess table size.
(OMP_HOST_VAR_CNT): Likewise.
(INT_MAX_STR_LEN): Constant for the maximal number of digits of a device
number.
(gomp_get_icv_flag): Returns if a flag for a particular ICV is set.
(gomp_set_icv_flag): Sets a flag for a particular ICV.
(print_device_specific_icvs): New helper function for omp_display_env to
print device specific ICV values.
(get_device_num): New helper function for parse_device_specific.
Extracts the device number from an environment variable name.
(get_icv_member_addr): Gets the memory address for a particular member
of an ICV struct.
(gomp_get_initial_icv_item): Get a list item of gomp_initial_icv_list.
(gomp_get_offload_icv_item): Get a list item of gomp_offload_icv_list.
(add_initial_icv_to_list): Adds an ICV struct to gomp_initial_icv_list.
(startswith): Checks if a string starts with a given prefix.
(initialize_env): Extended to parse the new syntax of environment
variables.
* icv-device.c (omp_get_max_teams): Added.
(ialias): Likewise.
(omp_set_num_teams): Likewise.
* icv.c (omp_set_num_teams): Moved to icv-device.c.
(omp_get_max_teams): Likewise.
(ialias): Likewise.
* libgomp-plugin.h (GOMP_DEVICE_NUM_VAR): Removed.
(GOMP_ADDITIONAL_ICVS): New target-side struct that
holds the designated ICVs of the target device.
* libgomp.h (enum gomp_icvs): Collection of ICVs.
(enum gomp_env_suffix): Collection of possible suffixes of environment
variables.
(struct gomp_initial_icvs): Contains all ICVs for which we need to store
initial values.
(struct gomp_default_icv_t): New struct to hold ICVs for which we need
to store initial values.
(struct gomp_icv_list): Definition of a linked list that is used for
storing ICVs for the devices and also for _DEV, _ALL, and without
suffix.
(gomp_get_initial_icv_item): Get a list item of gomp_initial_icv_list.
(gomp_get_icv_flag): Returns if a flag for a particular ICV is set.
(struct gomp_offload_icvs): New struct to hold ICVs that are copied to
a device.
(struct gomp_offload_icv_list): Definition of a linked list that holds
device-specific ICVs that are copied to devices.
(gomp_get_offload_icv_item): Get a list item of gomp_offload_icv_list.
* plugin/plugin-gcn.c (GOMP_OFFLOAD_load_image): Extended to read
further ICVs from the offload image.
* plugin/plugin-nvptx.c (GOMP_OFFLOAD_load_image): Likewise.
* target.c (get_gomp_offload_icvs): New. Returns the ICV values
depending on the device num and the variable hierarchy.
(gomp_load_image_to_device): Extended to copy further ICVs to a device.
* testsuite/libgomp.c-c++-common/icv-5.c: New test.
* testsuite/libgomp.c-c++-common/icv-6.c: New test.
* testsuite/libgomp.c-c++-common/icv-7.c: New test.
* testsuite/libgomp.c-c++-common/icv-8.c: New test.
* testsuite/libgomp.c-c++-common/omp-display-env-1.c: New test.
* testsuite/libgomp.c-c++-common/omp-display-env-2.c: New test.
diff --git a/libgomp/config/gcn/icv-device.c b/libgomp/config/gcn/icv-device.c
index f70b7e6..bf757ba 100644
--- a/libgomp/config/gcn/icv-device.c
+++ b/libgomp/config/gcn/icv-device.c
@@ -28,6 +28,10 @@
#include "libgomp.h"
+/* This is set to the ICV values of current GPU during device initialization,
+ when the offload image containing this libgomp portion is loaded. */
+volatile struct gomp_offload_icvs GOMP_ADDITIONAL_ICVS;
+
void
omp_set_default_device (int device_num __attribute__((unused)))
{
@@ -36,7 +40,7 @@ omp_set_default_device (int device_num __attribute__((unused)))
int
omp_get_default_device (void)
{
- return 0;
+ return GOMP_ADDITIONAL_ICVS.default_device;
}
int
@@ -58,14 +62,23 @@ omp_is_initial_device (void)
return 0;
}
-/* This is set to the device number of current GPU during device initialization,
- when the offload image containing this libgomp portion is loaded. */
-volatile int GOMP_DEVICE_NUM_VAR;
-
int
omp_get_device_num (void)
{
- return GOMP_DEVICE_NUM_VAR;
+ return GOMP_ADDITIONAL_ICVS.device_num;
+}
+
+int
+omp_get_max_teams (void)
+{
+ return GOMP_ADDITIONAL_ICVS.nteams;
+}
+
+void
+omp_set_num_teams (int num_teams)
+{
+ if (num_teams >= 0)
+ GOMP_ADDITIONAL_ICVS.nteams = num_teams;
}
ialias (omp_set_default_device)
@@ -74,3 +87,5 @@ ialias (omp_get_initial_device)
ialias (omp_get_num_devices)
ialias (omp_is_initial_device)
ialias (omp_get_device_num)
+ialias (omp_get_max_teams)
+ialias (omp_set_num_teams)
diff --git a/libgomp/config/nvptx/icv-device.c b/libgomp/config/nvptx/icv-device.c
index faf90f9..6f869be 100644
--- a/libgomp/config/nvptx/icv-device.c
+++ b/libgomp/config/nvptx/icv-device.c
@@ -28,6 +28,10 @@
#include "libgomp.h"
+/* This is set to the ICV values of current GPU during device initialization,
+ when the offload image containing this libgomp portion is loaded. */
+static volatile struct gomp_offload_icvs GOMP_ADDITIONAL_ICVS;
+
void
omp_set_default_device (int device_num __attribute__((unused)))
{
@@ -36,7 +40,7 @@ omp_set_default_device (int device_num __attribute__((unused)))
int
omp_get_default_device (void)
{
- return 0;
+ return GOMP_ADDITIONAL_ICVS.default_device;
}
int
@@ -58,14 +62,23 @@ omp_is_initial_device (void)
return 0;
}
-/* This is set to the device number of current GPU during device initialization,
- when the offload image containing this libgomp portion is loaded. */
-static volatile int GOMP_DEVICE_NUM_VAR;
-
int
omp_get_device_num (void)
{
- return GOMP_DEVICE_NUM_VAR;
+ return GOMP_ADDITIONAL_ICVS.device_num;
+}
+
+int
+omp_get_max_teams (void)
+{
+ return GOMP_ADDITIONAL_ICVS.nteams;
+}
+
+void
+omp_set_num_teams (int num_teams)
+{
+ if (num_teams >= 0)
+ GOMP_ADDITIONAL_ICVS.nteams = num_teams;
}
ialias (omp_set_default_device)
@@ -74,3 +87,5 @@ ialias (omp_get_initial_device)
ialias (omp_get_num_devices)
ialias (omp_is_initial_device)
ialias (omp_get_device_num)
+ialias (omp_get_max_teams)
+ialias (omp_set_num_teams)
diff --git a/libgomp/env.c b/libgomp/env.c
index 1c4ee89..9654dd3 100644
--- a/libgomp/env.c
+++ b/libgomp/env.c
@@ -73,6 +73,25 @@ struct gomp_task_icv gomp_global_icv = {
.target_data = NULL
};
+/* List for initial "_DEV", "_ALL", and "_DEV_X" ICVs like OMP_NUM_TEAMS_DEV,
+ OMP_NUM_TEAMS_ALL, or OMP_NUM_TEAMS_DEV_42. */
+struct gomp_icv_list *gomp_initial_icv_list = NULL;
+
+/* List for "_DEV_X" ICVs like OMP_NUM_TEAMS_DEV_42. This list contains all
+ device-specific ICVs that are copied from host to device and back. */
+struct gomp_offload_icv_list *gomp_offload_icv_list = NULL;
+
+/* Default values of ICVs according to the OpenMP standard. */
+struct gomp_default_icv_t gomp_default_icv_values = {
+ .run_sched_var = GFS_DYNAMIC,
+ .run_sched_chunk_size = 1,
+ .max_active_levels_var = 1,
+ .bind_var = omp_proc_bind_false,
+ .nteams_var = 0,
+ .teams_thread_limit_var = 0,
+ .default_device_var = 0
+};
+
bool gomp_cancel_var = false;
enum gomp_target_offload_t gomp_target_offload_var
= GOMP_TARGET_OFFLOAD_DEFAULT;
@@ -104,86 +123,94 @@ int goacc_default_dims[GOMP_DIM_MAX];
static int wait_policy;
static unsigned long stacksize = GOMP_DEFAULT_STACKSIZE;
-/* Parse the OMP_SCHEDULE environment variable. */
-
static void
-parse_schedule (void)
+print_env_var_error (const char *env, const char *val)
{
- char *env, *end;
+ char name[val - env];
+ memcpy (name, env, val - env - 1);
+ name[val - env - 1] = '\0';
+ gomp_error ("Invalid value for environment variable %s: %s", name, val);
+}
+
+/* Parse the OMP_SCHEDULE environment variable. */
+static bool
+parse_schedule (const char *env, const char *val, void * const params[])
+{
+ enum gomp_schedule_type *schedule = (enum gomp_schedule_type *) params[0];
+ int *chunk_size = (int *) params[1];
+ char *end;
unsigned long value;
int monotonic = 0;
- env = getenv ("OMP_SCHEDULE");
- if (env == NULL)
- return;
+ if (val == NULL)
+ return false;
- while (isspace ((unsigned char) *env))
- ++env;
- if (strncasecmp (env, "monotonic", 9) == 0)
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (strncasecmp (val, "monotonic", 9) == 0)
{
monotonic = 1;
- env += 9;
+ val += 9;
}
- else if (strncasecmp (env, "nonmonotonic", 12) == 0)
+ else if (strncasecmp (val, "nonmonotonic", 12) == 0)
{
monotonic = -1;
- env += 12;
+ val += 12;
}
if (monotonic)
{
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env != ':')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val != ':')
goto unknown;
- ++env;
- while (isspace ((unsigned char) *env))
- ++env;
+ ++val;
+ while (isspace ((unsigned char) *val))
+ ++val;
}
- if (strncasecmp (env, "static", 6) == 0)
+ if (strncasecmp (val, "static", 6) == 0)
{
- gomp_global_icv.run_sched_var = GFS_STATIC;
- env += 6;
+ *schedule = GFS_STATIC;
+ val += 6;
}
- else if (strncasecmp (env, "dynamic", 7) == 0)
+ else if (strncasecmp (val, "dynamic", 7) == 0)
{
- gomp_global_icv.run_sched_var = GFS_DYNAMIC;
- env += 7;
+ *schedule = GFS_DYNAMIC;
+ val += 7;
}
- else if (strncasecmp (env, "guided", 6) == 0)
+ else if (strncasecmp (val, "guided", 6) == 0)
{
- gomp_global_icv.run_sched_var = GFS_GUIDED;
- env += 6;
+ *schedule = GFS_GUIDED;
+ val += 6;
}
- else if (strncasecmp (env, "auto", 4) == 0)
+ else if (strncasecmp (val, "auto", 4) == 0)
{
- gomp_global_icv.run_sched_var = GFS_AUTO;
- env += 4;
+ *schedule = GFS_AUTO;
+ val += 4;
}
else
goto unknown;
if (monotonic == 1
- || (monotonic == 0 && gomp_global_icv.run_sched_var == GFS_STATIC))
- gomp_global_icv.run_sched_var |= GFS_MONOTONIC;
+ || (monotonic == 0 && *schedule == GFS_STATIC))
+ *schedule |= GFS_MONOTONIC;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
{
- gomp_global_icv.run_sched_chunk_size
- = (gomp_global_icv.run_sched_var & ~GFS_MONOTONIC) != GFS_STATIC;
- return;
+ *chunk_size = (*schedule & ~GFS_MONOTONIC) != GFS_STATIC;
+ return true;
}
- if (*env++ != ',')
+ if (*val++ != ',')
goto unknown;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
goto invalid;
errno = 0;
- value = strtoul (env, &end, 10);
- if (errno || end == env)
+ value = strtoul (val, &end, 10);
+ if (errno || end == val)
goto invalid;
while (isspace ((unsigned char) *end))
@@ -194,20 +221,22 @@ parse_schedule (void)
if ((int)value != value)
goto invalid;
- if (value == 0
- && (gomp_global_icv.run_sched_var & ~GFS_MONOTONIC) != GFS_STATIC)
+ if (value == 0 && (*schedule & ~GFS_MONOTONIC) != GFS_STATIC)
value = 1;
- gomp_global_icv.run_sched_chunk_size = value;
- return;
+ *chunk_size = value;
+ return true;
unknown:
- gomp_error ("Unknown value for environment variable OMP_SCHEDULE");
- return;
+ print_env_var_error (env, val);
+ return false;
invalid:
+ char name[val - env];
+ memcpy (name, env, val - env - 1);
+ name[val - env - 1] = '\0';
gomp_error ("Invalid value for chunk size in "
- "environment variable OMP_SCHEDULE");
- return;
+ "environment variable %s: %s", name, val);
+ return false;
}
/* Parse an unsigned long environment variable. Return true if one was
@@ -215,24 +244,23 @@ parse_schedule (void)
environment variable. */
static bool
-parse_unsigned_long_1 (const char *name, unsigned long *pvalue, bool allow_zero,
- bool secure)
+parse_unsigned_long_1 (const char *env, const char *val, unsigned long *pvalue,
+ bool allow_zero)
{
- char *env, *end;
+ char *end;
unsigned long value;
- env = (secure ? secure_getenv (name) : getenv (name));
- if (env == NULL)
+ if (val == NULL)
return false;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
goto invalid;
errno = 0;
- value = strtoul (env, &end, 10);
- if (errno || end == env || (long) value <= 0 - allow_zero)
+ value = strtoul (val, &end, 10);
+ if (errno || end == val || (long) value <= 0 - allow_zero)
goto invalid;
while (isspace ((unsigned char) *end))
@@ -244,16 +272,36 @@ parse_unsigned_long_1 (const char *name, unsigned long *pvalue, bool allow_zero,
return true;
invalid:
- gomp_error ("Invalid value for environment variable %s", name);
+ print_env_var_error (env, val);
return false;
}
/* As parse_unsigned_long_1, but always use getenv. */
static bool
-parse_unsigned_long (const char *name, unsigned long *pvalue, bool allow_zero)
+parse_unsigned_long (const char *env, const char *val, void * const params[])
{
- return parse_unsigned_long_1 (name, pvalue, allow_zero, false);
+ unsigned upper = (unsigned long) params[2];
+ unsigned long pvalue = 0;
+ bool ret = parse_unsigned_long_1 (env, val, &pvalue, (bool) params[1]);
+ if (!ret)
+ return false;
+
+ if (upper == 0)
+ *(unsigned long *) params[0] = pvalue;
+ else
+ {
+ if (pvalue > upper)
+ pvalue = upper;
+ if (upper <= UCHAR_MAX)
+ *(unsigned char *) params[0] = pvalue;
+ else if (upper <= UINT_MAX)
+ *(unsigned int *) params[0] = pvalue;
+ else
+ *(unsigned long *) params[0] = pvalue;
+ }
+
+ return ret;
}
/* Parse a positive int environment variable. Return true if one was
@@ -261,58 +309,57 @@ parse_unsigned_long (const char *name, unsigned long *pvalue, bool allow_zero)
environment variable. */
static bool
-parse_int_1 (const char *name, int *pvalue, bool allow_zero, bool secure)
+parse_int_1 (const char *env, const char *val, int *pvalue, bool allow_zero)
{
unsigned long value;
- if (!parse_unsigned_long_1 (name, &value, allow_zero, secure))
+ if (!parse_unsigned_long_1 (env, val, &value, allow_zero))
return false;
if (value > INT_MAX)
{
- gomp_error ("Invalid value for environment variable %s", name);
+ print_env_var_error (env, val);
return false;
}
*pvalue = (int) value;
return true;
}
-/* As parse_int_1, but use getenv. */
-
static bool
-parse_int (const char *name, int *pvalue, bool allow_zero)
+parse_int (const char *env, const char *val, void * const params[])
{
- return parse_int_1 (name, pvalue, allow_zero, false);
+ return parse_int_1 (env, val, (int *) params[0], (bool) params[1]);
}
/* As parse_int_1, but use getenv_secure. */
static bool
-parse_int_secure (const char *name, int *pvalue, bool allow_zero)
+parse_int_secure (const char *env, int *pvalue, bool allow_zero)
{
- return parse_int_1 (name, pvalue, allow_zero, true);
+ return parse_int_1 (env, secure_getenv (env), pvalue, allow_zero);
}
/* Parse an unsigned long list environment variable. Return true if one was
present and it was successfully parsed. */
static bool
-parse_unsigned_long_list (const char *name, unsigned long *p1stvalue,
- unsigned long **pvalues,
- unsigned long *pnvalues)
+parse_unsigned_long_list (const char *env, const char *val,
+ void * const params[])
{
- char *env, *end;
+ unsigned long *p1stvalue = (unsigned long *) params[0];
+ unsigned long **pvalues = (unsigned long **) params[1];
+ unsigned long *pnvalues = (unsigned long*) params[2];
+ char *end;
unsigned long value, *values = NULL;
- env = getenv (name);
- if (env == NULL)
+ if (val == NULL)
return false;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
goto invalid;
errno = 0;
- value = strtoul (env, &end, 10);
+ value = strtoul (val, &end, 10);
if (errno || (long) value <= 0)
goto invalid;
@@ -326,7 +373,7 @@ parse_unsigned_long_list (const char *name, unsigned long *p1stvalue,
do
{
- env = end + 1;
+ val = end + 1;
if (nvalues == nalloced)
{
unsigned long *n;
@@ -335,6 +382,9 @@ parse_unsigned_long_list (const char *name, unsigned long *p1stvalue,
if (n == NULL)
{
free (values);
+ char name[val - env];
+ memcpy (name, env, val - env - 1);
+ name[val - env - 1] = '\0';
gomp_error ("Out of memory while trying to parse"
" environment variable %s", name);
return false;
@@ -344,13 +394,13 @@ parse_unsigned_long_list (const char *name, unsigned long *p1stvalue,
values[nvalues++] = value;
}
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
goto invalid;
errno = 0;
- value = strtoul (env, &end, 10);
+ value = strtoul (val, &end, 10);
if (errno || (long) value <= 0)
goto invalid;
@@ -370,52 +420,56 @@ parse_unsigned_long_list (const char *name, unsigned long *p1stvalue,
}
goto invalid;
}
+ else
+ {
+ *pnvalues = 0;
+ *pvalues = NULL;
+ }
*p1stvalue = value;
return true;
invalid:
free (values);
- gomp_error ("Invalid value for environment variable %s", name);
+ print_env_var_error (env, val);
return false;
}
-static void
-parse_target_offload (const char *name, enum gomp_target_offload_t *offload)
+static bool
+parse_target_offload (const char *env, const char *val, void * const params[])
{
- const char *env;
int new_offload = -1;
- env = getenv (name);
- if (env == NULL)
- return;
+ if (val == NULL)
+ return false;
- while (isspace ((unsigned char) *env))
- ++env;
- if (strncasecmp (env, "default", 7) == 0)
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (strncasecmp (val, "default", 7) == 0)
{
- env += 7;
+ val += 7;
new_offload = GOMP_TARGET_OFFLOAD_DEFAULT;
}
- else if (strncasecmp (env, "mandatory", 9) == 0)
+ else if (strncasecmp (val, "mandatory", 9) == 0)
{
- env += 9;
+ val += 9;
new_offload = GOMP_TARGET_OFFLOAD_MANDATORY;
}
- else if (strncasecmp (env, "disabled", 8) == 0)
+ else if (strncasecmp (val, "disabled", 8) == 0)
{
- env += 8;
+ val += 8;
new_offload = GOMP_TARGET_OFFLOAD_DISABLED;
}
- while (isspace ((unsigned char) *env))
- ++env;
- if (new_offload != -1 && *env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (new_offload != -1 && *val == '\0')
{
- *offload = new_offload;
- return;
+ *(enum gomp_target_offload_t *) params[0] = new_offload;
+ return true;
}
- gomp_error ("Invalid value for environment variable OMP_TARGET_OFFLOAD");
+ print_env_var_error (env, val);
+ return false;
}
/* Parse environment variable set to a boolean or list of omp_proc_bind_t
@@ -423,10 +477,11 @@ parse_target_offload (const char *name, enum gomp_target_offload_t *offload)
parsed. */
static bool
-parse_bind_var (const char *name, char *p1stvalue,
- char **pvalues, unsigned long *pnvalues)
+parse_bind_var (const char *env, const char *val, void * const params[])
{
- char *env;
+ char *p1stvalue = (char *) params[0];
+ char **pvalues = (char **) params[1];
+ unsigned long *pnvalues = (unsigned long *) params[2];
char value = omp_proc_bind_false, *values = NULL;
int i;
static struct proc_bind_kinds
@@ -444,30 +499,29 @@ parse_bind_var (const char *name, char *p1stvalue,
{ "spread", 6, omp_proc_bind_spread }
};
- env = getenv (name);
- if (env == NULL)
+ if (val == NULL)
return false;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
goto invalid;
for (i = 0; i < 6; i++)
- if (strncasecmp (env, kinds[i].name, kinds[i].len) == 0)
+ if (strncasecmp (val, kinds[i].name, kinds[i].len) == 0)
{
value = kinds[i].kind;
- env += kinds[i].len;
+ val += kinds[i].len;
break;
}
if (i == 6)
goto invalid;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env != '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val != '\0')
{
- if (*env == ',')
+ if (*val == ',')
{
unsigned long nvalues = 0, nalloced = 0;
@@ -477,7 +531,7 @@ parse_bind_var (const char *name, char *p1stvalue,
do
{
- env++;
+ val++;
if (nvalues == nalloced)
{
char *n;
@@ -486,6 +540,9 @@ parse_bind_var (const char *name, char *p1stvalue,
if (n == NULL)
{
free (values);
+ char name[val - env];
+ memcpy (name, env, val - env - 1);
+ name[val - env - 1] = '\0';
gomp_error ("Out of memory while trying to parse"
" environment variable %s", name);
return false;
@@ -495,27 +552,27 @@ parse_bind_var (const char *name, char *p1stvalue,
values[nvalues++] = value;
}
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
goto invalid;
for (i = 2; i < 6; i++)
- if (strncasecmp (env, kinds[i].name, kinds[i].len) == 0)
+ if (strncasecmp (val, kinds[i].name, kinds[i].len) == 0)
{
value = kinds[i].kind;
- env += kinds[i].len;
+ val += kinds[i].len;
break;
}
if (i == 6)
goto invalid;
values[nvalues++] = value;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
break;
- if (*env != ',')
+ if (*val != ',')
goto invalid;
}
while (1);
@@ -532,7 +589,7 @@ parse_bind_var (const char *name, char *p1stvalue,
invalid:
free (values);
- gomp_error ("Invalid value for environment variable %s", name);
+ print_env_var_error (env, val);
return false;
}
@@ -865,23 +922,22 @@ parse_places_var (const char *name, bool ignore)
present and it was successfully parsed. */
static bool
-parse_stacksize (const char *name, unsigned long *pvalue)
+parse_stacksize (const char *env, const char *val, void * const params[])
{
- char *env, *end;
+ char *end;
unsigned long value, shift = 10;
- env = getenv (name);
- if (env == NULL)
+ if (val == NULL)
return false;
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
goto invalid;
errno = 0;
- value = strtoul (env, &end, 10);
- if (errno || end == env)
+ value = strtoul (val, &end, 10);
+ if (errno || end == val)
goto invalid;
while (isspace ((unsigned char) *end))
@@ -914,11 +970,11 @@ parse_stacksize (const char *name, unsigned long *pvalue)
if (((value << shift) >> shift) != value)
goto invalid;
- *pvalue = value << shift;
+ *(unsigned long *) params[0] = value << shift;
return true;
invalid:
- gomp_error ("Invalid value for environment variable %s", name);
+ print_env_var_error (env, val);
return false;
}
@@ -998,35 +1054,33 @@ parse_spincount (const char *name, unsigned long long *pvalue)
/* Parse a boolean value for environment variable NAME and store the
result in VALUE. Return true if one was present and it was
successfully parsed. */
-
static bool
-parse_boolean (const char *name, bool *value)
+parse_boolean (const char *env, const char *val, void * const params[])
{
- const char *env;
+ bool *value = (bool *) params[0];
- env = getenv (name);
- if (env == NULL)
+ if (val == NULL)
return false;
- while (isspace ((unsigned char) *env))
- ++env;
- if (strncasecmp (env, "true", 4) == 0)
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (strncasecmp (val, "true", 4) == 0)
{
*value = true;
- env += 4;
+ val += 4;
}
- else if (strncasecmp (env, "false", 5) == 0)
+ else if (strncasecmp (val, "false", 5) == 0)
{
*value = false;
- env += 5;
+ val += 5;
}
else
- env = "X";
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env != '\0')
+ val = "X";
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val != '\0')
{
- gomp_error ("Invalid value for environment variable %s", name);
+ print_env_var_error (env, val);
return false;
}
return true;
@@ -1034,36 +1088,42 @@ parse_boolean (const char *name, bool *value)
/* Parse the OMP_WAIT_POLICY environment variable and return the value. */
-static int
-parse_wait_policy (void)
+static bool
+parse_wait_policy (const char *env, const char *val, void * const params[])
{
- const char *env;
+ int *pvalue = (int *) params[0];
int ret = -1;
- env = getenv ("OMP_WAIT_POLICY");
- if (env == NULL)
- return -1;
+ if (val == NULL)
+ {
+ *pvalue = -1;
+ return false;
+ }
- while (isspace ((unsigned char) *env))
- ++env;
- if (strncasecmp (env, "active", 6) == 0)
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (strncasecmp (val, "active", 6) == 0)
{
ret = 1;
- env += 6;
+ val += 6;
}
- else if (strncasecmp (env, "passive", 7) == 0)
+ else if (strncasecmp (val, "passive", 7) == 0)
{
ret = 0;
- env += 7;
+ val += 7;
}
else
- env = "X";
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
- return ret;
- gomp_error ("Invalid value for environment variable OMP_WAIT_POLICY");
- return -1;
+ val = "X";
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
+ {
+ *pvalue = ret;
+ return true;
+ }
+ print_env_var_error (env, val);
+ *pvalue = -1;
+ return false;
}
/* Parse the GOMP_CPU_AFFINITY environment varible. Return true if one was
@@ -1167,26 +1227,24 @@ parse_affinity (bool ignore)
}
/* Parse the OMP_ALLOCATOR environment variable and return the value. */
-
-static uintptr_t
-parse_allocator (void)
+static bool
+parse_allocator (const char *env, const char *val, void * const params[])
{
- const char *env;
- uintptr_t ret = omp_default_mem_alloc;
+ uintptr_t *ret = (uintptr_t *) params[0];
+ *ret = omp_default_mem_alloc;
- env = getenv ("OMP_ALLOCATOR");
- if (env == NULL)
- return ret;
+ if (val == NULL)
+ return false;
- while (isspace ((unsigned char) *env))
- ++env;
+ while (isspace ((unsigned char) *val))
+ ++val;
if (0)
;
#define C(v) \
- else if (strncasecmp (env, #v, sizeof (#v) - 1) == 0) \
+ else if (strncasecmp (val, #v, sizeof (#v) - 1) == 0) \
{ \
- ret = v; \
- env += sizeof (#v) - 1; \
+ *ret = v; \
+ val += sizeof (#v) - 1; \
}
C (omp_default_mem_alloc)
C (omp_large_cap_mem_alloc)
@@ -1198,13 +1256,14 @@ parse_allocator (void)
C (omp_thread_mem_alloc)
#undef C
else
- env = "X";
- while (isspace ((unsigned char) *env))
- ++env;
- if (*env == '\0')
- return ret;
- gomp_error ("Invalid value for environment variable OMP_ALLOCATOR");
- return omp_default_mem_alloc;
+ val = "X";
+ while (isspace ((unsigned char) *val))
+ ++val;
+ if (*val == '\0')
+ return true;
+ print_env_var_error (env, val);
+ *ret = omp_default_mem_alloc;
+ return false;
}
static void
@@ -1251,62 +1310,59 @@ parse_gomp_openacc_dim (void)
}
}
-void
-omp_display_env (int verbose)
-{
- int i;
+/* Helper function for omp_display_env which prints the values of run_sched_var.
+ 'device' can be 'host', 'dev', 'all' or a particular device number. */
- fputs ("\nOPENMP DISPLAY ENVIRONMENT BEGIN\n", stderr);
-
- fputs (" _OPENMP = '201511'\n", stderr);
- fprintf (stderr, " OMP_DYNAMIC = '%s'\n",
- gomp_global_icv.dyn_var ? "TRUE" : "FALSE");
- fprintf (stderr, " OMP_NESTED = '%s'\n",
- gomp_global_icv.max_active_levels_var > 1 ? "TRUE" : "FALSE");
-
- fprintf (stderr, " OMP_NUM_THREADS = '%lu", gomp_global_icv.nthreads_var);
- for (i = 1; i < gomp_nthreads_var_list_len; i++)
- fprintf (stderr, ",%lu", gomp_nthreads_var_list[i]);
- fputs ("'\n", stderr);
-
- fprintf (stderr, " OMP_SCHEDULE = '");
- if ((gomp_global_icv.run_sched_var & GFS_MONOTONIC))
+static void
+print_schedule (enum gomp_schedule_type run_sched_var, int run_sched_chunk_size,
+ const char *device)
+{
+ fprintf (stderr, " [%s] OMP_SCHEDULE = '", device);
+ if ((run_sched_var & GFS_MONOTONIC))
{
- if (gomp_global_icv.run_sched_var != (GFS_MONOTONIC | GFS_STATIC))
+ if (run_sched_var != (GFS_MONOTONIC | GFS_STATIC))
fputs ("MONOTONIC:", stderr);
}
- else if (gomp_global_icv.run_sched_var == GFS_STATIC)
+ else if (run_sched_var == GFS_STATIC)
fputs ("NONMONOTONIC:", stderr);
- switch (gomp_global_icv.run_sched_var & ~GFS_MONOTONIC)
+ switch (run_sched_var & ~GFS_MONOTONIC)
{
case GFS_RUNTIME:
fputs ("RUNTIME", stderr);
- if (gomp_global_icv.run_sched_chunk_size != 1)
- fprintf (stderr, ",%d", gomp_global_icv.run_sched_chunk_size);
+ if (run_sched_chunk_size != 1)
+ fprintf (stderr, ",%d", run_sched_chunk_size);
break;
case GFS_STATIC:
fputs ("STATIC", stderr);
- if (gomp_global_icv.run_sched_chunk_size != 0)
- fprintf (stderr, ",%d", gomp_global_icv.run_sched_chunk_size);
+ if (run_sched_chunk_size != 0)
+ fprintf (stderr, ",%d", run_sched_chunk_size);
break;
case GFS_DYNAMIC:
fputs ("DYNAMIC", stderr);
- if (gomp_global_icv.run_sched_chunk_size != 1)
- fprintf (stderr, ",%d", gomp_global_icv.run_sched_chunk_size);
+ if (run_sched_chunk_size != 1)
+ fprintf (stderr, ",%d", run_sched_chunk_size);
break;
case GFS_GUIDED:
fputs ("GUIDED", stderr);
- if (gomp_global_icv.run_sched_chunk_size != 1)
- fprintf (stderr, ",%d", gomp_global_icv.run_sched_chunk_size);
+ if (run_sched_chunk_size != 1)
+ fprintf (stderr, ",%d", run_sched_chunk_size);
break;
case GFS_AUTO:
fputs ("AUTO", stderr);
break;
}
fputs ("'\n", stderr);
+}
+
+/* Helper function for omp_display_env which prints the values of proc_bind_var.
+ 'device' can be 'host', 'dev', 'all', or a particular device number. */
- fputs (" OMP_PROC_BIND = '", stderr);
- switch (gomp_global_icv.bind_var)
+static void
+print_proc_bind (char proc_bind_var, unsigned long len, char **list,
+ const char *device)
+{
+ fprintf (stderr, " [%s] OMP_PROC_BIND = '", device);
+ switch (proc_bind_var)
{
case omp_proc_bind_false:
fputs ("FALSE", stderr);
@@ -1324,8 +1380,8 @@ omp_display_env (int verbose)
fputs ("SPREAD", stderr);
break;
}
- for (i = 1; i < gomp_bind_var_list_len; i++)
- switch (gomp_bind_var_list[i])
+ for (int i = 1; i < len; i++)
+ switch ((*list)[i])
{
case omp_proc_bind_master:
fputs (",MASTER", stderr); /* TODO: Change to PRIMARY for OpenMP 5.1. */
@@ -1338,7 +1394,284 @@ omp_display_env (int verbose)
break;
}
fputs ("'\n", stderr);
- fputs (" OMP_PLACES = '", stderr);
+}
+
+enum gomp_parse_type
+{
+ PARSE_INT = 1,
+ PARSE_BOOL = 2,
+ PARSE_UINT = 3,
+ PARSE_ULONG = 4,
+ PARSE_CHAR = 5,
+ PARSE_UCHAR = 6
+};
+
+/* The following table contains items that help parsing environment variables
+ and fill corresponding ICVs with values. FLAG_VARS contain all ICVS which
+ are affected by the environment variable. FLAGS determine what variant of
+ environment variable is allowed. */
+
+#define ENTRY(NAME) NAME, sizeof (NAME) - 1
+static const struct envvar
+{
+ const char *name;
+ int name_len;
+ uint8_t flag_vars[3];
+ uint8_t flag;
+ bool (*parse_func) (const char *, const char *, void * const[]);
+} envvars[] = {
+ { ENTRY ("OMP_SCHEDULE"),
+ { GOMP_ICV_SCHEDULE, GOMP_ICV_SCHEDULE_CHUNK_SIZE },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_schedule },
+ { ENTRY ("OMP_NUM_TEAMS"),
+ { GOMP_ICV_NTEAMS },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_int },
+ { ENTRY ("OMP_DYNAMIC"),
+ { GOMP_ICV_DYNAMIC },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_boolean },
+ { ENTRY ("OMP_TEAMS_THREAD_LIMIT"),
+ { GOMP_ICV_TEAMS_THREAD_LIMIT },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_int },
+ { ENTRY ("OMP_THREAD_LIMIT"),
+ { GOMP_ICV_THREAD_LIMIT },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_unsigned_long },
+ { ENTRY ("OMP_NUM_THREADS"),
+ { GOMP_ICV_NTHREADS, GOMP_ICV_NTHREADS_LIST, GOMP_ICV_NTHREADS_LIST_LEN },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_unsigned_long_list },
+ { ENTRY ("OMP_PROC_BIND"),
+ { GOMP_ICV_BIND, GOMP_ICV_BIND_LIST, GOMP_ICV_BIND_LIST_LEN },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_bind_var },
+ { ENTRY ("OMP_MAX_ACTIVE_LEVELS"),
+ { GOMP_ICV_MAX_ACTIVE_LEVELS },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_unsigned_long },
+ { ENTRY ("OMP_WAIT_POLICY"),
+ { GOMP_ICV_WAIT_POLICY },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_wait_policy },
+ { ENTRY ("OMP_STACKSIZE"),
+ { GOMP_ICV_STACKSIZE },
+ GOMP_ENV_SUFFIX_DEV | GOMP_ENV_SUFFIX_ALL | GOMP_ENV_SUFFIX_DEV_X,
+ &parse_stacksize },
+ { ENTRY ("OMP_CANCELLATION"), { GOMP_ICV_CANCELLATION }, 0, &parse_boolean },
+ { ENTRY ("OMP_DISPLAY_AFFINITY"), { GOMP_ICV_DISPLAY_AFFINITY }, 0,
+ &parse_boolean },
+ { ENTRY ("OMP_TARGET_OFFLOAD"), { GOMP_ICV_TARGET_OFFLOAD }, 0,
+ &parse_target_offload },
+ { ENTRY ("OMP_MAX_TASK_PRIORITY"), { GOMP_ICV_MAX_TASK_PRIORITY }, 0,
+ &parse_int },
+ { ENTRY ("OMP_ALLOCATOR"), { GOMP_ICV_ALLOCATOR }, 0, &parse_allocator },
+ { ENTRY ("OMP_DEFAULT_DEVICE"), { GOMP_ICV_DEFAULT_DEVICE }, 0, &parse_int }
+};
+#undef ENTRY
+#define OMP_VAR_CNT (sizeof (envvars) / sizeof (envvars[0]))
+
+/* The following table is used to apply the hierarchy of ICV variants for host
+ variables, e.g. nteams_var is set to OMP_NUM_TEAMS_ALL if OMP_NUM_TEAMS is
+ undefined. */
+
+static const struct host_envvar
+{
+ unsigned char flag_var;
+ void *dest[3];
+ int type_code;
+} host_envvars[] = {
+ { GOMP_ICV_NTEAMS, { &gomp_nteams_var }, PARSE_INT },
+ { GOMP_ICV_DYNAMIC, { &gomp_global_icv.dyn_var }, PARSE_BOOL },
+ { GOMP_ICV_DEFAULT_DEVICE, { &gomp_global_icv.default_device_var },
+ PARSE_INT },
+ { GOMP_ICV_TEAMS_THREAD_LIMIT, { &gomp_teams_thread_limit_var }, PARSE_INT },
+ { GOMP_ICV_SCHEDULE,
+ { &gomp_global_icv.run_sched_var, &gomp_global_icv.run_sched_chunk_size },
+ PARSE_INT },
+ { GOMP_ICV_THREAD_LIMIT, { &gomp_global_icv.thread_limit_var }, PARSE_UINT },
+ { GOMP_ICV_NTHREADS,
+ { &gomp_global_icv.nthreads_var, &gomp_nthreads_var_list,
+ &gomp_nthreads_var_list_len }, PARSE_ULONG },
+ { GOMP_ICV_BIND,
+ { &gomp_global_icv.bind_var, &gomp_bind_var_list, &gomp_bind_var_list_len },
+ PARSE_CHAR },
+ { GOMP_ICV_MAX_ACTIVE_LEVELS, { &gomp_global_icv.max_active_levels_var },
+ PARSE_UCHAR },
+};
+#define OMP_HOST_VAR_CNT (sizeof (host_envvars) / sizeof (host_envvars[0]))
+
+#define INT_MAX_STR_LEN 10
+
+bool
+gomp_get_icv_flag (uint32_t value, enum gomp_icvs icv)
+{
+ return value & (1 << (icv - 1));
+}
+
+static void
+gomp_set_icv_flag (uint32_t *value, enum gomp_icvs icv)
+{
+ *value |= 1 << (icv - 1);
+}
+
+static void
+print_device_specific_icvs (int icv_code)
+{
+ struct gomp_icv_list *list = gomp_initial_icv_list;
+ int i;
+ char dev_num[INT_MAX_STR_LEN + 1];
+
+ while (list != NULL)
+ {
+ if (list->device_num < 0)
+ {
+ list = list->next;
+ continue;
+ }
+
+ switch (icv_code)
+ {
+ case GOMP_ICV_NTEAMS:
+ if (gomp_get_icv_flag (list->flags, GOMP_ICV_NTEAMS))
+ fprintf (stderr, " [%d] OMP_NUM_TEAMS = '%d'\n",
+ list->device_num, list->icvs.nteams_var);
+ break;
+ case GOMP_ICV_DYNAMIC:
+ if (gomp_get_icv_flag (list->flags, GOMP_ICV_DYNAMIC))
+ fprintf (stderr, " [%d] OMP_DYNAMIC = '%s'\n",
+ list->device_num, list->icvs.dyn_var ? "TRUE" : "FALSE");
+ break;
+ case GOMP_ICV_TEAMS_THREAD_LIMIT:
+ if (gomp_get_icv_flag (list->flags, GOMP_ICV_TEAMS_THREAD_LIMIT))
+ fprintf (stderr, " [%d] OMP_TEAMS_THREAD_LIMIT = '%u'\n",
+ list->device_num, list->icvs.teams_thread_limit_var);
+ break;
+ case GOMP_ICV_SCHEDULE:
+ if (!(gomp_get_icv_flag (list->flags, GOMP_ICV_SCHEDULE)))
+ break;
+ sprintf (dev_num, "%d", list->device_num);
+ print_schedule (list->icvs.run_sched_var,
+ list->icvs.run_sched_chunk_size,
+ dev_num);
+ break;
+ case GOMP_ICV_THREAD_LIMIT:
+ if (gomp_get_icv_flag (list->flags, GOMP_ICV_THREAD_LIMIT))
+ fprintf (stderr, " [%d] OMP_THREAD_LIMIT = '%d'\n",
+ list->device_num, list->icvs.thread_limit_var);
+ break;
+ case GOMP_ICV_NTHREADS:
+ if (!(gomp_get_icv_flag (list->flags, GOMP_ICV_NTHREADS)))
+ break;
+ fprintf (stderr, " [%d] OMP_NUM_THREADS = '%lu", list->device_num,
+ list->icvs.nthreads_var);
+ for (i = 1; i < list->icvs.nthreads_var_list_len; i++)
+ fprintf (stderr, ",%lu", list->icvs.nthreads_var_list[i]);
+ fputs ("'\n", stderr);
+ break;
+ case GOMP_ICV_MAX_ACTIVE_LEVELS:
+ fprintf (stderr, " [%d] OMP_MAX_ACTIVE_LEVELS = '%u'\n",
+ list->device_num, list->icvs.max_active_levels_var);
+ break;
+ case GOMP_ICV_BIND:
+ if (!(gomp_get_icv_flag (list->flags, GOMP_ICV_BIND)))
+ break;
+ sprintf (dev_num, "%d", list->device_num);
+ print_proc_bind (list->icvs.bind_var, list->icvs.bind_var_list_len,
+ &list->icvs.bind_var_list, dev_num);
+ break;
+ case GOMP_ICV_WAIT_POLICY:
+ if (gomp_get_icv_flag (list->flags, GOMP_ICV_WAIT_POLICY))
+ fprintf (stderr, " [%d] OMP_WAIT_POLICY = '%s'\n",
+ list->device_num,
+ list->icvs.wait_policy > 0 ? "ACTIVE" : "PASSIVE");
+ break;
+ case GOMP_ICV_STACKSIZE:
+ if (gomp_get_icv_flag (list->flags, GOMP_ICV_STACKSIZE))
+ fprintf (stderr, " [%d] OMP_STACKSIZE = '%lu'\n",
+ list->device_num, list->icvs.stacksize);
+ break;
+ }
+ list = list->next;
+ }
+}
+
+void
+omp_display_env (int verbose)
+{
+ int i;
+ struct gomp_icv_list *dev = gomp_get_initial_icv_item (-1);
+ struct gomp_icv_list *all = gomp_get_initial_icv_item (-2);
+ struct gomp_icv_list *none = gomp_get_initial_icv_item (-3);
+
+ fputs ("\nOPENMP DISPLAY ENVIRONMENT BEGIN\n", stderr);
+
+ fputs (" _OPENMP = '201511'\n", stderr);
+
+ fprintf (stderr, " [host] OMP_DYNAMIC = '%s'\n",
+ none->icvs.dyn_var ? "TRUE" : "FALSE");
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_DYNAMIC))
+ fprintf (stderr, " [all] OMP_DYNAMIC = '%s'\n",
+ all->icvs.dyn_var ? "TRUE" : "FALSE");
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_DYNAMIC))
+ fprintf (stderr, " [device] OMP_DYNAMIC = '%s'\n",
+ dev->icvs.dyn_var ? "TRUE" : "FALSE");
+ print_device_specific_icvs (GOMP_ICV_DYNAMIC);
+
+ /* The OMP_NESTED environment variable has been deprecated. */
+ fprintf (stderr, " [host] OMP_NESTED = '%s'\n",
+ none->icvs.max_active_levels_var > 1 ? "TRUE" : "FALSE");
+
+ fprintf (stderr, " [host] OMP_NUM_THREADS = '%lu",
+ none->icvs.nthreads_var);
+ for (i = 1; i < none->icvs.nthreads_var_list_len; i++)
+ fprintf (stderr, ",%lu", none->icvs.nthreads_var_list[i]);
+ fputs ("'\n", stderr);
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_NTHREADS))
+ {
+ fprintf (stderr, " [all] OMP_NUM_THREADS = '%lu",
+ all->icvs.nthreads_var);
+ for (i = 1; i < all->icvs.nthreads_var_list_len; i++)
+ fprintf (stderr, ",%lu", all->icvs.nthreads_var_list[i]);
+ fputs ("'\n", stderr);
+ }
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_NTHREADS))
+ {
+ fprintf (stderr, " [device] OMP_NUM_THREADS = '%lu",
+ dev->icvs.nthreads_var);
+ for (i = 1; i < dev->icvs.nthreads_var_list_len; i++)
+ fprintf (stderr, ",%lu", dev->icvs.nthreads_var_list[i]);
+ fputs ("'\n", stderr);
+ }
+ print_device_specific_icvs (GOMP_ICV_NTHREADS);
+
+
+ print_schedule (none->icvs.run_sched_var,
+ none->icvs.run_sched_chunk_size, "host");
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_SCHEDULE))
+ print_schedule (all->icvs.run_sched_var,
+ all->icvs.run_sched_chunk_size, "all");
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_SCHEDULE))
+ print_schedule (dev->icvs.run_sched_var,
+ dev->icvs.run_sched_chunk_size, "device");
+ print_device_specific_icvs (GOMP_ICV_SCHEDULE);
+
+ print_proc_bind (none->icvs.bind_var,
+ none->icvs.bind_var_list_len,
+ &none->icvs.bind_var_list, "host");
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_BIND))
+ print_proc_bind (all->icvs.bind_var,
+ all->icvs.bind_var_list_len,
+ &all->icvs.bind_var_list, "all");
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_BIND))
+ print_proc_bind (dev->icvs.bind_var,
+ dev->icvs.bind_var_list_len,
+ &dev->icvs.bind_var_list, "device");
+ print_device_specific_icvs (GOMP_ICV_BIND);
+
+ fputs (" [host] OMP_PLACES = '", stderr);
for (i = 0; i < gomp_places_list_len; i++)
{
fputs ("{", stderr);
@@ -1347,30 +1680,85 @@ omp_display_env (int verbose)
}
fputs ("'\n", stderr);
- fprintf (stderr, " OMP_STACKSIZE = '%lu'\n", stacksize);
+ fprintf (stderr, " [host] OMP_STACKSIZE = '%lu'\n",
+ none->icvs.stacksize);
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_STACKSIZE))
+ fprintf (stderr, " [all] OMP_STACKSIZE = '%lu'\n",
+ all->icvs.stacksize);
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_STACKSIZE))
+ fprintf (stderr, " [device] OMP_STACKSIZE = '%lu'\n",
+ dev->icvs.stacksize);
+ print_device_specific_icvs (GOMP_ICV_STACKSIZE);
/* GOMP's default value is actually neither active nor passive. */
- fprintf (stderr, " OMP_WAIT_POLICY = '%s'\n",
- wait_policy > 0 ? "ACTIVE" : "PASSIVE");
- fprintf (stderr, " OMP_THREAD_LIMIT = '%u'\n",
- gomp_global_icv.thread_limit_var);
- fprintf (stderr, " OMP_MAX_ACTIVE_LEVELS = '%u'\n",
- gomp_global_icv.max_active_levels_var);
- fprintf (stderr, " OMP_NUM_TEAMS = '%u'\n", gomp_nteams_var);
- fprintf (stderr, " OMP_TEAMS_THREAD_LIMIT = '%u'\n",
- gomp_teams_thread_limit_var);
-
- fprintf (stderr, " OMP_CANCELLATION = '%s'\n",
+ fprintf (stderr, " [host] OMP_WAIT_POLICY = '%s'\n",
+ none->icvs.wait_policy > 0 ? "ACTIVE" : "PASSIVE");
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_WAIT_POLICY))
+ fprintf (stderr, " [all] OMP_WAIT_POLICY = '%s'\n",
+ all->icvs.wait_policy > 0 ? "ACTIVE" : "PASSIVE");
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_WAIT_POLICY))
+ fprintf (stderr, " [device] OMP_WAIT_POLICY = '%s'\n",
+ dev->icvs.wait_policy > 0 ? "ACTIVE" : "PASSIVE");
+ print_device_specific_icvs (GOMP_ICV_WAIT_POLICY);
+
+ fprintf (stderr, " [host] OMP_THREAD_LIMIT = '%u'\n",
+ none->icvs.thread_limit_var);
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_THREAD_LIMIT))
+ fprintf (stderr, " [all] OMP_THREAD_LIMIT = '%d'\n",
+ all->icvs.thread_limit_var);
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_THREAD_LIMIT))
+ fprintf (stderr, " [device] OMP_THREAD_LIMIT = '%d'\n",
+ dev->icvs.thread_limit_var);
+ print_device_specific_icvs (GOMP_ICV_THREAD_LIMIT);
+
+ fprintf (stderr, " [host] OMP_MAX_ACTIVE_LEVELS = '%u'\n",
+ none->icvs.max_active_levels_var);
+ if (all != NULL && gomp_get_icv_flag (all->flags,
+ GOMP_ICV_MAX_ACTIVE_LEVELS))
+ fprintf (stderr, " [all] OMP_MAX_ACTIVE_LEVELS = '%u'\n",
+ all->icvs.max_active_levels_var);
+ if (dev != NULL && gomp_get_icv_flag (dev->flags,
+ GOMP_ICV_MAX_ACTIVE_LEVELS))
+ fprintf (stderr, " [device] OMP_MAX_ACTIVE_LEVELS = '%u'\n",
+ dev->icvs.max_active_levels_var);
+ print_device_specific_icvs (GOMP_ICV_MAX_ACTIVE_LEVELS);
+
+
+ fprintf (stderr, " [host] OMP_NUM_TEAMS = '%d'\n",
+ none->icvs.nteams_var);
+ if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_NTEAMS))
+ fprintf (stderr, " [all] OMP_NUM_TEAMS = '%d'\n",
+ all->icvs.nteams_var);
+ if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_NTEAMS))
+ fprintf (stderr, " [device] OMP_NUM_TEAMS = '%d'\n",
+ dev->icvs.nteams_var);
+ print_device_specific_icvs (GOMP_ICV_NTEAMS);
+
+ fprintf (stderr, " [host] OMP_TEAMS_THREAD_LIMIT = '%u'\n",
+ none->icvs.teams_thread_limit_var);
+ if (all != NULL && gomp_get_icv_flag (all->flags,
+ GOMP_ICV_TEAMS_THREAD_LIMIT))
+ fprintf (stderr, " [all] OMP_TEAMS_THREAD_LIMIT = '%u'\n",
+ all->icvs.teams_thread_limit_var);
+ if (dev != NULL && gomp_get_icv_flag (dev->flags,
+ GOMP_ICV_TEAMS_THREAD_LIMIT))
+ fprintf (stderr, " [device] OMP_TEAMS_THREAD_LIMIT = '%u'\n",
+ dev->icvs.teams_thread_limit_var);
+ print_device_specific_icvs (GOMP_ICV_TEAMS_THREAD_LIMIT);
+
+ fprintf (stderr, " [host] OMP_CANCELLATION = '%s'\n",
gomp_cancel_var ? "TRUE" : "FALSE");
- fprintf (stderr, " OMP_DEFAULT_DEVICE = '%d'\n",
- gomp_global_icv.default_device_var);
- fprintf (stderr, " OMP_MAX_TASK_PRIORITY = '%d'\n",
+
+ fprintf (stderr, " [host] OMP_DEFAULT_DEVICE = '%d'\n",
+ none->icvs.default_device_var);
+
+ fprintf (stderr, " [host] OMP_MAX_TASK_PRIORITY = '%d'\n",
gomp_max_task_priority_var);
- fprintf (stderr, " OMP_DISPLAY_AFFINITY = '%s'\n",
+ fprintf (stderr, " [host] OMP_DISPLAY_AFFINITY = '%s'\n",
gomp_display_affinity_var ? "TRUE" : "FALSE");
- fprintf (stderr, " OMP_AFFINITY_FORMAT = '%s'\n",
+ fprintf (stderr, " [host] OMP_AFFINITY_FORMAT = '%s'\n",
gomp_affinity_format_var);
- fprintf (stderr, " OMP_ALLOCATOR = '");
+ fprintf (stderr, " [host] OMP_ALLOCATOR = '");
switch (gomp_def_allocator)
{
#define C(v) case v: fputs (#v, stderr); break;
@@ -1387,7 +1775,7 @@ omp_display_env (int verbose)
}
fputs ("'\n", stderr);
- fputs (" OMP_TARGET_OFFLOAD = '", stderr);
+ fputs (" [host] OMP_TARGET_OFFLOAD = '", stderr);
switch (gomp_target_offload_var)
{
case GOMP_TARGET_OFFLOAD_DEFAULT:
@@ -1404,13 +1792,13 @@ omp_display_env (int verbose)
if (verbose)
{
- fputs (" GOMP_CPU_AFFINITY = ''\n", stderr);
- fprintf (stderr, " GOMP_STACKSIZE = '%lu'\n", stacksize);
+ fputs (" [host] GOMP_CPU_AFFINITY = ''\n", stderr);
+ fprintf (stderr, " [host] GOMP_STACKSIZE = '%lu'\n", stacksize);
#ifdef HAVE_INTTYPES_H
- fprintf (stderr, " GOMP_SPINCOUNT = '%"PRIu64"'\n",
+ fprintf (stderr, " [host] GOMP_SPINCOUNT = '%"PRIu64"'\n",
(uint64_t) gomp_spin_count_var);
#else
- fprintf (stderr, " GOMP_SPINCOUNT = '%lu'\n",
+ fprintf (stderr, " [host] GOMP_SPINCOUNT = '%lu'\n",
(unsigned long) gomp_spin_count_var);
#endif
}
@@ -1459,65 +1847,381 @@ handle_omp_display_env (void)
ialias_call (omp_display_env) (verbose);
}
+/* Helper function for initialize_env. Extracts the device number from
+ an environment variable name. ENV is the complete environment variable.
+ DEV_NUM_PTR points to the start of the device number in the environment
+ variable string. DEV_NUM_LEN is the returned length of the device num
+ string. */
+
+static bool
+get_device_num (char *env, char *dev_num_ptr, int *dev_num, int *dev_num_len)
+{
+ char *end;
+ int pos = 0;
+
+ if (dev_num_ptr[0] == '-')
+ {
+ gomp_error ("Non-negative device number expected in %s", env);
+ return false;
+ }
+
+ while (pos <= INT_MAX_STR_LEN)
+ {
+ if (dev_num_ptr[pos] == '\0' || dev_num_ptr[pos] == '=')
+ break;
+ pos++;
+ }
+ if (pos > INT_MAX_STR_LEN)
+ {
+ gomp_error ("Invalid device number in %s (too long)", env);
+ return false;
+ }
+
+ *dev_num = (int) strtoul (dev_num_ptr, &end, 10);
+ if (dev_num_ptr[0] == '0' && *dev_num != 0)
+ {
+ gomp_error ("Invalid device number in %s (leading zero)", env);
+ return false;
+ }
+ if (dev_num_ptr == end || *end != '=')
+ {
+ gomp_error ("Invalid device number in %s", env);
+ return false;
+ }
+
+ *dev_num_len = pos;
+ return true;
+}
+
+static void
+get_icv_member_addr (struct gomp_initial_icvs *icvs, int icv_code,
+ void *icv_addr[3])
+{
+ if (icv_code == 0 || icv_addr == NULL)
+ return;
+
+ icv_addr[0] = icv_addr[1] = icv_addr[2] = NULL;
+
+ switch (icv_code)
+ {
+ case GOMP_ICV_NTEAMS:
+ icv_addr[0] = &(icvs->nteams_var);
+ icv_addr[1] = false;
+ break;
+ case GOMP_ICV_DYNAMIC:
+ icv_addr[0] = &((*icvs).dyn_var);
+ break;
+ case GOMP_ICV_TEAMS_THREAD_LIMIT:
+ icv_addr[0] = &(icvs->teams_thread_limit_var);
+ icv_addr[1] = false;
+ break;
+ case GOMP_ICV_SCHEDULE:
+ icv_addr[0] = &(icvs->run_sched_var);
+ icv_addr[1] = &(icvs->run_sched_chunk_size);
+ break;
+ case GOMP_ICV_THREAD_LIMIT:
+ icv_addr[0] = &(icvs->thread_limit_var);
+ icv_addr[1] = false;
+ icv_addr[2] = (void *) UINT_MAX;
+ break;
+ case GOMP_ICV_NTHREADS:
+ icv_addr[0] = &(icvs->nthreads_var);
+ icv_addr[1] = &(icvs->nthreads_var_list);
+ icv_addr[2] = &(icvs->nthreads_var_list_len);
+ break;
+ case GOMP_ICV_MAX_ACTIVE_LEVELS:
+ icv_addr[0] = &(icvs->max_active_levels_var);
+ icv_addr[1] = (void *) true;
+ icv_addr[2] = (void *) gomp_supported_active_levels;
+ break;
+ case GOMP_ICV_BIND:
+ icv_addr[0] = &(icvs->bind_var);
+ icv_addr[1] = &(icvs->bind_var_list);
+ icv_addr[2] = &(icvs->bind_var_list_len);
+ break;
+ case GOMP_ICV_WAIT_POLICY:
+ icv_addr[0] = &(icvs->wait_policy);
+ break;
+ case GOMP_ICV_STACKSIZE:
+ icv_addr[0] = &(icvs->stacksize);
+ break;
+ case GOMP_ICV_CANCELLATION:
+ icv_addr[0] = &gomp_cancel_var;
+ break;
+ case GOMP_ICV_DISPLAY_AFFINITY:
+ icv_addr[0] = &gomp_display_affinity_var;
+ break;
+ case GOMP_ICV_TARGET_OFFLOAD:
+ icv_addr[0] = &gomp_target_offload_var;
+ break;
+ case GOMP_ICV_MAX_TASK_PRIORITY:
+ icv_addr[0] = &gomp_max_task_priority_var;
+ break;
+ case GOMP_ICV_ALLOCATOR:
+ icv_addr[0] = &gomp_def_allocator;
+ break;
+ case GOMP_ICV_DEFAULT_DEVICE:
+ icv_addr[0] = &(icvs->default_device_var);
+ icv_addr[1] = (void *) true;
+ break;
+ }
+}
+
+struct gomp_icv_list *
+gomp_get_initial_icv_item (int dev_num)
+{
+ struct gomp_icv_list *l = gomp_initial_icv_list;
+ while (l != NULL && l->device_num != dev_num)
+ l = l->next;
+
+ return l;
+}
+
+struct gomp_offload_icv_list *
+gomp_get_offload_icv_item (int dev_num)
+{
+ struct gomp_offload_icv_list *l = gomp_offload_icv_list;
+ while (l != NULL && l->device_num != dev_num)
+ l = l->next;
+
+ return l;
+}
+
+/* Helper function for initialize_env to add a device specific ICV value
+ to gomp_initial_icv_list. */
+
+static uint32_t *
+add_initial_icv_to_list (int dev_num, int icv_code, void *icv_addr[3])
+{
+ struct gomp_icv_list *last = NULL, *l = gomp_initial_icv_list;
+ while (l != NULL && l->device_num != dev_num)
+ {
+ last = l;
+ l = l->next;
+ }
+
+ if (l == NULL)
+ {
+ l
+ = (struct gomp_icv_list *) gomp_malloc (sizeof (struct gomp_icv_list));
+ l->device_num = dev_num;
+ memset (&(l->icvs), 0, sizeof (struct gomp_initial_icvs));
+ l->flags = 0;
+ if (dev_num < 0)
+ {
+ l->next = gomp_initial_icv_list;
+ gomp_initial_icv_list = l;
+ }
+ else
+ {
+ l->next = NULL;
+ if (last == NULL)
+ gomp_initial_icv_list = l;
+ else
+ last->next = l;
+ }
+ }
+
+ get_icv_member_addr (&(l->icvs), icv_code, icv_addr);
+
+ return &(l->flags);
+}
+
+/* Return true if STR string starts with PREFIX. */
+
+static inline bool
+startswith (const char *str, const char *prefix)
+{
+ return strncmp (str, prefix, strlen (prefix)) == 0;
+}
static void __attribute__((constructor))
initialize_env (void)
{
- unsigned long thread_limit_var;
- unsigned long max_active_levels_var;
+ extern char **environ;
+ char **env;
+ int omp_var, dev_num = 0, dev_num_len = 0, i;
+ bool ignore = false;
+ char *env_val;
+ void *params[3];
+ uint32_t *flag_var_addr = NULL;
+ unsigned pos;
+ struct gomp_icv_list *all, *none;
/* Do a compile time check that mkomp_h.pl did good job. */
omp_check_defines ();
- parse_schedule ();
- parse_boolean ("OMP_DYNAMIC", &gomp_global_icv.dyn_var);
- parse_boolean ("OMP_CANCELLATION", &gomp_cancel_var);
- parse_boolean ("OMP_DISPLAY_AFFINITY", &gomp_display_affinity_var);
- parse_int ("OMP_DEFAULT_DEVICE", &gomp_global_icv.default_device_var, true);
- parse_target_offload ("OMP_TARGET_OFFLOAD", &gomp_target_offload_var);
- parse_int ("OMP_MAX_TASK_PRIORITY", &gomp_max_task_priority_var, true);
- gomp_def_allocator = parse_allocator ();
- if (parse_unsigned_long ("OMP_THREAD_LIMIT", &thread_limit_var, false))
- {
- gomp_global_icv.thread_limit_var
- = thread_limit_var > INT_MAX ? UINT_MAX : thread_limit_var;
- }
- parse_int_secure ("GOMP_DEBUG", &gomp_debug_var, true);
#ifndef HAVE_SYNC_BUILTINS
gomp_mutex_init (&gomp_managed_threads_lock);
#endif
gomp_init_num_threads ();
gomp_available_cpus = gomp_global_icv.nthreads_var;
- if (!parse_unsigned_long_list ("OMP_NUM_THREADS",
- &gomp_global_icv.nthreads_var,
- &gomp_nthreads_var_list,
- &gomp_nthreads_var_list_len))
- gomp_global_icv.nthreads_var = gomp_available_cpus;
- parse_int ("OMP_NUM_TEAMS", &gomp_nteams_var, false);
- parse_int ("OMP_TEAMS_THREAD_LIMIT", &gomp_teams_thread_limit_var, false);
- bool ignore = false;
- if (parse_bind_var ("OMP_PROC_BIND",
- &gomp_global_icv.bind_var,
- &gomp_bind_var_list,
- &gomp_bind_var_list_len)
+
+ /* Initial values for host environment variables should always exist even if
+ there is no explicitly set host environment variable. Moreover, they are
+ set to the initial global values. */
+ add_initial_icv_to_list (-3, 0, NULL);
+ none = gomp_get_initial_icv_item (-3);
+ none->icvs.nthreads_var = 1;
+ none->icvs.thread_limit_var = UINT_MAX;
+ none->icvs.run_sched_var = GFS_DYNAMIC;
+ none->icvs.run_sched_chunk_size = 1;
+ none->icvs.default_device_var = 0;
+ none->icvs.dyn_var = false;
+ none->icvs.max_active_levels_var = 1;
+ none->icvs.bind_var = omp_proc_bind_false;
+
+ for (env = environ; *env != 0; env++)
+ {
+ if (!startswith (*env, "OMP_"))
+ continue;
+
+ for (omp_var = 0; omp_var < OMP_VAR_CNT; omp_var++)
+ {
+ if (startswith (*env, envvars[omp_var].name))
+ {
+ pos = envvars[omp_var].name_len;
+ if ((*env)[pos] == '=')
+ {
+ pos++;
+ flag_var_addr
+ = add_initial_icv_to_list (-3,
+ envvars[omp_var].flag_vars[0],
+ params);
+ }
+ else if (startswith (&(*env)[pos], "_DEV=")
+ && envvars[omp_var].flag & GOMP_ENV_SUFFIX_DEV)
+ {
+ pos += 5;
+ flag_var_addr
+ = add_initial_icv_to_list (-1,
+ envvars[omp_var].flag_vars[0],
+ params);
+ }
+ else if (startswith (&(*env)[pos], "_ALL=")
+ && envvars[omp_var].flag & GOMP_ENV_SUFFIX_ALL)
+ {
+ pos += 5;
+ flag_var_addr
+ = add_initial_icv_to_list (-2,
+ envvars[omp_var].flag_vars[0],
+ params);
+ }
+ else if (startswith (&(*env)[pos], "_DEV_")
+ && envvars[omp_var].flag & GOMP_ENV_SUFFIX_DEV_X)
+ {
+ pos += 5;
+ if (!get_device_num (*env, &(*env)[pos], &dev_num,
+ &dev_num_len))
+ goto next_var;
+
+ pos += dev_num_len + 1;
+ flag_var_addr
+ = add_initial_icv_to_list (dev_num,
+ envvars[omp_var].flag_vars[0],
+ params);
+ }
+ else
+ {
+ gomp_error ("Invalid device number in %s", *env);
+ break;
+ }
+ env_val = &(*env)[pos];
+
+ if (envvars[omp_var].parse_func (*env, env_val, params))
+ {
+ for (i = 0; i < 3; ++i)
+ if (envvars[omp_var].flag_vars[i])
+ gomp_set_icv_flag (flag_var_addr,
+ envvars[omp_var].flag_vars[i]);
+ else
+ break;
+ }
+
+ break;
+ }
+ }
+
+ next_var:
+ }
+
+ all = gomp_get_initial_icv_item (-2);
+ for (omp_var = 0; omp_var < OMP_HOST_VAR_CNT; omp_var++)
+ {
+ if (none != NULL
+ && gomp_get_icv_flag (none->flags, host_envvars[omp_var].flag_var))
+ get_icv_member_addr (&none->icvs,
+ host_envvars[omp_var].flag_var, params);
+ else if (all != NULL
+ && gomp_get_icv_flag (all->flags,
+ host_envvars[omp_var].flag_var))
+ get_icv_member_addr (&all->icvs, host_envvars[omp_var].flag_var,
+ params);
+ else
+ continue;
+
+ switch (host_envvars[omp_var].type_code)
+ {
+ case PARSE_INT:
+ for (i = 0; i < 3; ++i)
+ if (host_envvars[omp_var].dest[i] != NULL && params[i] != NULL)
+ *(int *) (host_envvars[omp_var].dest[i]) = *(int *) params[i];
+ break;
+ case PARSE_BOOL:
+ for (i = 0; i < 3; ++i)
+ if (host_envvars[omp_var].dest[i] != NULL && params[i] != NULL)
+ *(bool *) (host_envvars[omp_var].dest[i]) = *(bool *) params[i];
+ break;
+ case PARSE_UINT:
+ for (i = 0; i < 3; ++i)
+ if (host_envvars[omp_var].dest[i] != NULL && params[i] != NULL)
+ *(unsigned int *) (host_envvars[omp_var].dest[i])
+ = *(unsigned int *) params[i];
+ break;
+ case PARSE_ULONG:
+ for (i = 0; i < 3; ++i)
+ if (host_envvars[omp_var].dest[i] != NULL && params[i] != NULL)
+ *(unsigned long *) (host_envvars[omp_var].dest[i])
+ = *(unsigned long *) params[i];
+ break;
+ case PARSE_CHAR:
+ for (i = 0; i < 3; ++i)
+ if (host_envvars[omp_var].dest[i] != NULL && params[i] != NULL)
+ *(char *) (host_envvars[omp_var].dest[i]) = *(char *) params[i];
+ break;
+ case PARSE_UCHAR:
+ for (i = 0; i < 3; ++i)
+ if (host_envvars[omp_var].dest[i] != NULL && params[i] != NULL)
+ *(unsigned char *) (host_envvars[omp_var].dest[i])
+ = *(unsigned char *) params[i];
+ break;
+ }
+ }
+
+ if (((none != NULL && gomp_get_icv_flag (none->flags, GOMP_ICV_BIND))
+ || (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_BIND)))
&& gomp_global_icv.bind_var == omp_proc_bind_false)
ignore = true;
- if (parse_unsigned_long ("OMP_MAX_ACTIVE_LEVELS",
- &max_active_levels_var, true))
- gomp_global_icv.max_active_levels_var
- = (max_active_levels_var > gomp_supported_active_levels)
- ? gomp_supported_active_levels : max_active_levels_var;
- else
+
+ if (!((none != NULL
+ && gomp_get_icv_flag (none->flags, GOMP_ICV_MAX_ACTIVE_LEVELS))
+ || (all != NULL
+ && gomp_get_icv_flag (all->flags, GOMP_ICV_MAX_ACTIVE_LEVELS))))
{
bool nested = true;
+ const char *env = getenv ("OMP_NESTED");
/* OMP_NESTED is deprecated in OpenMP 5.0. */
- if (parse_boolean ("OMP_NESTED", &nested))
+ if (parse_boolean ("OMP_NESTED", env, (void *[]) {&nested}))
gomp_global_icv.max_active_levels_var
= nested ? gomp_supported_active_levels : 1;
else if (gomp_nthreads_var_list_len > 1 || gomp_bind_var_list_len > 1)
gomp_global_icv.max_active_levels_var = gomp_supported_active_levels;
}
+
+ /* Process GOMP_* variables and dependencies between parsed ICVs. */
+ parse_int_secure ("GOMP_DEBUG", &gomp_debug_var, true);
+
/* Make sure OMP_PLACES and GOMP_CPU_AFFINITY env vars are always
parsed if present in the environment. If OMP_PROC_BIND was set
explicitly to false, don't populate places list though. If places
@@ -1547,7 +2251,11 @@ initialize_env (void)
gomp_set_affinity_format (env, strlen (env));
}
- wait_policy = parse_wait_policy ();
+ if (none != NULL && gomp_get_icv_flag (none->flags, GOMP_ICV_WAIT_POLICY))
+ wait_policy = none->icvs.wait_policy;
+ else if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_WAIT_POLICY))
+ wait_policy = all->icvs.wait_policy;
+
if (!parse_spincount ("GOMP_SPINCOUNT", &gomp_spin_count_var))
{
/* Using a rough estimation of 100000 spins per msec,
@@ -1573,8 +2281,21 @@ initialize_env (void)
/* Not strictly environment related, but ordering constructors is tricky. */
pthread_attr_init (&gomp_thread_attr);
- if (parse_stacksize ("OMP_STACKSIZE", &stacksize)
- || parse_stacksize ("GOMP_STACKSIZE", &stacksize)
+ if (!(none != NULL && gomp_get_icv_flag (none->flags, GOMP_ICV_STACKSIZE)))
+ {
+ const char *env = getenv ("GOMP_STACKSIZE");
+ if (env != NULL
+ && parse_stacksize ("GOMP_STACKSIZE", env,
+ (void *[3]) {&none->icvs.stacksize}))
+ gomp_set_icv_flag (&none->flags, GOMP_ICV_STACKSIZE);
+ }
+ if (none != NULL && gomp_get_icv_flag (none->flags, GOMP_ICV_STACKSIZE))
+ stacksize = none->icvs.stacksize;
+ else if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_STACKSIZE))
+ stacksize = all->icvs.stacksize;
+
+ if ((none != NULL && gomp_get_icv_flag (none->flags, GOMP_ICV_STACKSIZE))
+ || (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_STACKSIZE))
|| GOMP_DEFAULT_STACKSIZE)
{
int err;
@@ -1601,7 +2322,8 @@ initialize_env (void)
/* OpenACC. */
- if (!parse_int ("ACC_DEVICE_NUM", &goacc_device_num, true))
+ if (!parse_int ("ACC_DEVICE_NUM", getenv ("ACC_DEVICE_NUM"),
+ (void *[]) {&goacc_device_num, (void *) true}))
goacc_device_num = 0;
parse_acc_device_type ();
diff --git a/libgomp/icv-device.c b/libgomp/icv-device.c
index 11ceb30..d8acf0e 100644
--- a/libgomp/icv-device.c
+++ b/libgomp/icv-device.c
@@ -80,3 +80,20 @@ omp_get_device_num (void)
}
ialias (omp_get_device_num)
+
+int
+omp_get_max_teams (void)
+{
+ return gomp_nteams_var;
+}
+
+ialias (omp_get_max_teams)
+
+void
+omp_set_num_teams (int num_teams)
+{
+ if (num_teams >= 0)
+ gomp_nteams_var = num_teams;
+}
+
+ialias (omp_set_num_teams)
diff --git a/libgomp/icv.c b/libgomp/icv.c
index de15cc8..df423c0 100644
--- a/libgomp/icv.c
+++ b/libgomp/icv.c
@@ -149,19 +149,6 @@ omp_get_supported_active_levels (void)
}
void
-omp_set_num_teams (int num_teams)
-{
- if (num_teams >= 0)
- gomp_nteams_var = num_teams;
-}
-
-int
-omp_get_max_teams (void)
-{
- return gomp_nteams_var;
-}
-
-void
omp_set_teams_thread_limit (int thread_limit)
{
if (thread_limit >= 0)
@@ -274,8 +261,6 @@ ialias (omp_get_thread_limit)
ialias (omp_set_max_active_levels)
ialias (omp_get_max_active_levels)
ialias (omp_get_supported_active_levels)
-ialias (omp_set_num_teams)
-ialias (omp_get_max_teams)
ialias (omp_set_teams_thread_limit)
ialias (omp_get_teams_thread_limit)
ialias (omp_get_cancellation)
diff --git a/libgomp/libgomp-plugin.h b/libgomp/libgomp-plugin.h
index ab3ed63..71a307f 100644
--- a/libgomp/libgomp-plugin.h
+++ b/libgomp/libgomp-plugin.h
@@ -102,11 +102,11 @@ struct addr_pair
uintptr_t end;
};
-/* This symbol is to name a target side variable that holds the designated
- 'device number' of the target device. The symbol needs to be available to
- libgomp code and the offload plugin (which in the latter case must be
- stringified). */
-#define GOMP_DEVICE_NUM_VAR __gomp_device_num
+/* This following symbol is used to name the target side variable struct that
+ holds the designated ICVs of the target device. The symbol needs to be
+ available to libgomp code and the offload plugin (which in the latter case
+ must be stringified). */
+#define GOMP_ADDITIONAL_ICVS __gomp_additional_icvs
/* Miscellaneous functions. */
extern void *GOMP_PLUGIN_malloc (size_t) __attribute__ ((malloc));
diff --git a/libgomp/libgomp.h b/libgomp/libgomp.h
index c243c4d..84e85f1 100644
--- a/libgomp/libgomp.h
+++ b/libgomp/libgomp.h
@@ -453,6 +453,31 @@ struct gomp_team_state
struct target_mem_desc;
+enum gomp_icvs
+{
+ GOMP_ICV_NTEAMS = 1,
+ GOMP_ICV_SCHEDULE = 2,
+ GOMP_ICV_SCHEDULE_CHUNK_SIZE = 3,
+ GOMP_ICV_DYNAMIC = 4,
+ GOMP_ICV_TEAMS_THREAD_LIMIT = 5,
+ GOMP_ICV_THREAD_LIMIT = 6,
+ GOMP_ICV_NTHREADS = 7,
+ GOMP_ICV_NTHREADS_LIST = 8,
+ GOMP_ICV_NTHREADS_LIST_LEN = 9,
+ GOMP_ICV_BIND = 10,
+ GOMP_ICV_BIND_LIST = 11,
+ GOMP_ICV_BIND_LIST_LEN = 12,
+ GOMP_ICV_MAX_ACTIVE_LEVELS = 13,
+ GOMP_ICV_WAIT_POLICY = 14,
+ GOMP_ICV_STACKSIZE = 15,
+ GOMP_ICV_DEFAULT_DEVICE = 16,
+ GOMP_ICV_CANCELLATION = 17,
+ GOMP_ICV_DISPLAY_AFFINITY = 18,
+ GOMP_ICV_TARGET_OFFLOAD = 19,
+ GOMP_ICV_MAX_TASK_PRIORITY = 20,
+ GOMP_ICV_ALLOCATOR = 21
+};
+
/* These are the OpenMP 4.0 Internal Control Variables described in
section 2.3.1. Those described as having one copy per task are
stored within the structure; those described as having one copy
@@ -472,6 +497,86 @@ struct gomp_task_icv
struct target_mem_desc *target_data;
};
+enum gomp_env_suffix
+{
+ GOMP_ENV_SUFFIX_UNKNOWN = 0,
+ GOMP_ENV_SUFFIX_NONE = 1,
+ GOMP_ENV_SUFFIX_DEV = 2,
+ GOMP_ENV_SUFFIX_ALL = 4,
+ GOMP_ENV_SUFFIX_DEV_X = 8
+};
+
+/* Struct that contains all ICVs for which we need to store initial values.
+ Keeping the initial values is needed for omp_display_env. Moreover initial
+ _DEV and _ALL variants of environment variables are also used to determine
+ actually used values for devices and for the host. */
+struct gomp_initial_icvs
+{
+ unsigned long *nthreads_var_list;
+ char *bind_var_list;
+ unsigned long nthreads_var;
+ unsigned long nthreads_var_list_len;
+ unsigned long bind_var_list_len;
+ unsigned long stacksize;
+ int run_sched_chunk_size;
+ int default_device_var;
+ int nteams_var;
+ int teams_thread_limit_var;
+ int wait_policy;
+ unsigned int thread_limit_var;
+ enum gomp_schedule_type run_sched_var;
+ bool dyn_var;
+ unsigned char max_active_levels_var;
+ char bind_var;
+};
+
+struct gomp_default_icv_t
+{
+ enum gomp_schedule_type run_sched_var;
+ int run_sched_chunk_size;
+ unsigned char max_active_levels_var;
+ char bind_var;
+ int nteams_var;
+ int teams_thread_limit_var;
+ int default_device_var;
+};
+extern struct gomp_default_icv_t gomp_default_icv_values;
+
+/* DEVICE_NUM "-1" is reserved for "_DEV" icvs.
+ DEVICE_NUM "-2" is reserved for "_ALL" icvs.
+ DEVICE_NUM "-3" is reserved for ICVs without suffix.
+ Non-negative DEVICE_NUM is for "_DEV_X" icvs. */
+struct gomp_icv_list
+{
+ int device_num;
+ struct gomp_initial_icvs icvs;
+ uint32_t flags;
+ struct gomp_icv_list *next;
+};
+extern struct gomp_icv_list *gomp_initial_icv_list;
+extern struct gomp_icv_list *gomp_get_initial_icv_item (int dev_num);
+
+extern bool gomp_get_icv_flag (uint32_t value, enum gomp_icvs icv);
+extern struct gomp_icv_list* gomp_add_device_specific_icv
+ (int dev_num, size_t size, struct gomp_icv_list **list);
+
+struct gomp_offload_icvs
+{
+ int device_num;
+ int default_device;
+ int nteams;
+ int teams_thread_limit;
+};
+
+struct gomp_offload_icv_list
+{
+ int device_num;
+ struct gomp_offload_icvs icvs;
+ struct gomp_offload_icv_list *next;
+};
+extern struct gomp_offload_icv_list *gomp_offload_icv_list;
+extern struct gomp_offload_icv_list *gomp_get_offload_icv_item (int dev_num);
+
enum gomp_target_offload_t
{
GOMP_TARGET_OFFLOAD_DEFAULT,
diff --git a/libgomp/plugin/plugin-gcn.c b/libgomp/plugin/plugin-gcn.c
index ea327bf..8821870 100644
--- a/libgomp/plugin/plugin-gcn.c
+++ b/libgomp/plugin/plugin-gcn.c
@@ -3367,6 +3367,7 @@ GOMP_OFFLOAD_load_image (int ord, unsigned version, const void *target_data,
struct kernel_info *kernel;
int kernel_count = image_desc->kernel_count;
unsigned var_count = image_desc->global_variable_count;
+ /* Currently, "others" is a struct of ICVS. */
int other_count = 1;
agent = get_agent_info (ord);
@@ -3464,36 +3465,39 @@ GOMP_OFFLOAD_load_image (int ord, unsigned version, const void *target_data,
}
}
- GCN_DEBUG ("Looking for variable %s\n", XSTRING (GOMP_DEVICE_NUM_VAR));
+ GCN_DEBUG ("Looking for variable %s\n", XSTRING (GOMP_ADDITIONAL_ICVS));
hsa_status_t status;
hsa_executable_symbol_t var_symbol;
status = hsa_fns.hsa_executable_get_symbol_fn (agent->executable, NULL,
- XSTRING (GOMP_DEVICE_NUM_VAR),
+ XSTRING (GOMP_ADDITIONAL_ICVS),
agent->id, 0, &var_symbol);
if (status == HSA_STATUS_SUCCESS)
{
- uint64_t device_num_varptr;
- uint32_t device_num_varsize;
+ uint64_t varptr;
+ uint32_t varsize;
status = hsa_fns.hsa_executable_symbol_get_info_fn
(var_symbol, HSA_EXECUTABLE_SYMBOL_INFO_VARIABLE_ADDRESS,
- &device_num_varptr);
+ &varptr);
if (status != HSA_STATUS_SUCCESS)
hsa_fatal ("Could not extract a variable from its symbol", status);
status = hsa_fns.hsa_executable_symbol_get_info_fn
(var_symbol, HSA_EXECUTABLE_SYMBOL_INFO_VARIABLE_SIZE,
- &device_num_varsize);
+ &varsize);
if (status != HSA_STATUS_SUCCESS)
- hsa_fatal ("Could not extract a variable size from its symbol", status);
+ hsa_fatal ("Could not extract a variable size from its symbol",
+ status);
- pair->start = device_num_varptr;
- pair->end = device_num_varptr + device_num_varsize;
+ pair->start = varptr;
+ pair->end = varptr + varsize;
}
else
- /* The 'GOMP_DEVICE_NUM_VAR' variable was not in this image. */
- pair->start = pair->end = 0;
- pair++;
+ {
+ /* The variable was not in this image. */
+ GCN_DEBUG ("Variable not found in image: %s\n", XSTRING (GOMP_ADDITIONAL_ICVS));
+ pair->start = pair->end = 0;
+ }
/* Ensure that constructors are run first. */
struct GOMP_kernel_launch_attributes kla =
diff --git a/libgomp/plugin/plugin-nvptx.c b/libgomp/plugin/plugin-nvptx.c
index bc63e274..a12f1ac 100644
--- a/libgomp/plugin/plugin-nvptx.c
+++ b/libgomp/plugin/plugin-nvptx.c
@@ -1305,7 +1305,7 @@ GOMP_OFFLOAD_load_image (int ord, unsigned version, const void *target_data,
fn_entries = img_header->fn_num;
fn_descs = img_header->fn_descs;
- /* Currently, the only other entry kind is 'device number'. */
+ /* Currently, other_entries contains only the struct of ICVs. */
other_entries = 1;
targ_tbl = GOMP_PLUGIN_malloc (sizeof (struct addr_pair)
@@ -1358,20 +1358,19 @@ GOMP_OFFLOAD_load_image (int ord, unsigned version, const void *target_data,
targ_tbl->end = targ_tbl->start + bytes;
}
- CUdeviceptr device_num_varptr;
- size_t device_num_varsize;
- CUresult r = CUDA_CALL_NOCHECK (cuModuleGetGlobal, &device_num_varptr,
- &device_num_varsize, module,
- XSTRING (GOMP_DEVICE_NUM_VAR));
+ CUdeviceptr varptr;
+ size_t varsize;
+ CUresult r = CUDA_CALL_NOCHECK (cuModuleGetGlobal, &varptr, &varsize,
+ module, XSTRING (GOMP_ADDITIONAL_ICVS));
+
if (r == CUDA_SUCCESS)
{
- targ_tbl->start = (uintptr_t) device_num_varptr;
- targ_tbl->end = (uintptr_t) (device_num_varptr + device_num_varsize);
+ targ_tbl->start = (uintptr_t) varptr;
+ targ_tbl->end = (uintptr_t) (varptr + varsize);
}
else
- /* The 'GOMP_DEVICE_NUM_VAR' variable was not in this image. */
+ /* The variable was not in this image. */
targ_tbl->start = targ_tbl->end = 0;
- targ_tbl++;
nvptx_set_clocktick (module, dev);
diff --git a/libgomp/target.c b/libgomp/target.c
index 135db1d..1624938 100644
--- a/libgomp/target.c
+++ b/libgomp/target.c
@@ -2108,6 +2108,56 @@ gomp_update (struct gomp_device_descr *devicep, size_t mapnum, void **hostaddrs,
gomp_mutex_unlock (&devicep->lock);
}
+/* Helper function for 'gomp_load_image_to_device'. Returns the ICV values
+ depending on the device num and the variable hierarchy
+ (_DEV_42, _DEV, _ALL). If no ICV was initially configured for the given
+ device and thus no item with that device number is contained in
+ gomp_offload_icv_list, then a new item is created and added to the list. */
+
+static struct gomp_offload_icvs *
+get_gomp_offload_icvs (int dev_num)
+{
+ struct gomp_icv_list *dev = gomp_get_initial_icv_item (-1);
+ struct gomp_icv_list *all = gomp_get_initial_icv_item (-2);
+ struct gomp_icv_list *dev_x = gomp_get_initial_icv_item (dev_num);
+ struct gomp_offload_icv_list *offload_icvs
+ = gomp_get_offload_icv_item (dev_num);
+
+ if (offload_icvs != NULL)
+ return &offload_icvs->icvs;
+
+ struct gomp_offload_icv_list *new
+ = (struct gomp_offload_icv_list *) gomp_malloc (sizeof (struct gomp_offload_icv_list));
+
+ new->device_num = dev_num;
+ new->icvs.device_num = dev_num;
+ new->next = gomp_offload_icv_list;
+
+ if (dev_x != NULL && gomp_get_icv_flag (dev_x->flags, GOMP_ICV_NTEAMS))
+ new->icvs.nteams = dev_x->icvs.nteams_var;
+ else if (dev != NULL && gomp_get_icv_flag (dev->flags, GOMP_ICV_NTEAMS))
+ new->icvs.nteams = dev->icvs.nteams_var;
+ else if (all != NULL && gomp_get_icv_flag (all->flags, GOMP_ICV_NTEAMS))
+ new->icvs.nteams = all->icvs.nteams_var;
+ else
+ new->icvs.nteams = gomp_default_icv_values.nteams_var;
+
+ if (dev_x != NULL
+ && gomp_get_icv_flag (dev_x->flags, GOMP_ICV_DEFAULT_DEVICE))
+ new->icvs.default_device = dev_x->icvs.default_device_var;
+ else if (dev != NULL
+ && gomp_get_icv_flag (dev->flags, GOMP_ICV_DEFAULT_DEVICE))
+ new->icvs.default_device = dev->icvs.default_device_var;
+ else if (all != NULL
+ && gomp_get_icv_flag (all->flags, GOMP_ICV_DEFAULT_DEVICE))
+ new->icvs.default_device = all->icvs.default_device_var;
+ else
+ new->icvs.default_device = gomp_default_icv_values.default_device_var;
+
+ gomp_offload_icv_list = new;
+ return &new->icvs;
+}
+
/* Load image pointed by TARGET_DATA to the device, specified by DEVICEP.
And insert to splay tree the mapping between addresses from HOST_TABLE and
from loaded target image. We rely in the host and device compiler
@@ -2128,9 +2178,6 @@ gomp_load_image_to_device (struct gomp_device_descr *devicep, unsigned version,
int num_funcs = host_funcs_end - host_func_table;
int num_vars = (host_vars_end - host_var_table) / 2;
- /* Others currently is only 'device_num' */
- int num_others = 1;
-
/* Load image to device and get target addresses for the image. */
struct addr_pair *target_table = NULL;
int i, num_target_entries;
@@ -2140,8 +2187,8 @@ gomp_load_image_to_device (struct gomp_device_descr *devicep, unsigned version,
target_data, &target_table);
if (num_target_entries != num_funcs + num_vars
- /* Others (device_num) are included as trailing entries in pair list. */
- && num_target_entries != num_funcs + num_vars + num_others)
+ /* "+1" due to the additional ICV struct. */
+ && num_target_entries != num_funcs + num_vars + 1)
{
gomp_mutex_unlock (&devicep->lock);
if (is_register_lock)
@@ -2153,7 +2200,9 @@ gomp_load_image_to_device (struct gomp_device_descr *devicep, unsigned version,
/* Insert host-target address mapping into splay tree. */
struct target_mem_desc *tgt = gomp_malloc (sizeof (*tgt));
- tgt->array = gomp_malloc ((num_funcs + num_vars) * sizeof (*tgt->array));
+ /* "+1" due to the additional ICV struct. */
+ tgt->array = gomp_malloc ((num_funcs + num_vars + 1)
+ * sizeof (*tgt->array));
tgt->refcount = REFCOUNT_INFINITY;
tgt->tgt_start = 0;
tgt->tgt_end = 0;
@@ -2213,32 +2262,40 @@ gomp_load_image_to_device (struct gomp_device_descr *devicep, unsigned version,
array++;
}
- /* Last entry is for the on-device 'device_num' variable. Tolerate case
- where plugin does not return this entry. */
+ /* Last entry is for a ICVs variable.
+ Tolerate case where plugin does not return those entries. */
if (num_funcs + num_vars < num_target_entries)
{
- struct addr_pair *device_num_var = &target_table[num_funcs + num_vars];
- /* Start address will be non-zero for last entry if GOMP_DEVICE_NUM_VAR
- was found in this image. */
- if (device_num_var->start != 0)
+ struct addr_pair *var = &target_table[num_funcs + num_vars];
+
+ /* Start address will be non-zero for the ICVs variable if
+ the variable was found in this image. */
+ if (var->start != 0)
{
/* The index of the devicep within devices[] is regarded as its
'device number', which is different from the per-device type
devicep->target_id. */
- int device_num_val = (int) (devicep - &devices[0]);
- if (device_num_var->end - device_num_var->start != sizeof (int))
- {
- gomp_mutex_unlock (&devicep->lock);
- if (is_register_lock)
- gomp_mutex_unlock (®ister_lock);
- gomp_fatal ("offload plugin managed 'device_num' not of expected "
- "format");
- }
-
- /* Copy device_num value to place on device memory, hereby actually
- designating its device number into effect. */
- gomp_copy_host2dev (devicep, NULL, (void *) device_num_var->start,
- &device_num_val, sizeof (int), false, NULL);
+ int dev_num = (int) (devicep - &devices[0]);
+ struct gomp_offload_icvs *icvs = get_gomp_offload_icvs (dev_num);
+ size_t var_size = var->end - var->start;
+
+ /* Copy the ICVs variable to place on device memory, hereby
+ actually designating its device number into effect. */
+ gomp_copy_host2dev (devicep, NULL, (void *) var->start, icvs,
+ var_size, false, NULL);
+ splay_tree_key k = &array->key;
+ k->host_start = (uintptr_t) icvs;
+ k->host_end =
+ k->host_start + (size_mask & sizeof (struct gomp_offload_icvs));
+ k->tgt = tgt;
+ k->tgt_offset = var->start;
+ k->refcount = REFCOUNT_INFINITY;
+ k->dynamic_refcount = 0;
+ k->aux = NULL;
+ array->left = NULL;
+ array->right = NULL;
+ splay_tree_insert (&devicep->mem_map, array);
+ array++;
}
}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/icv-5.c b/libgomp/testsuite/libgomp.c-c++-common/icv-5.c
new file mode 100644
index 0000000..82108bce
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/icv-5.c
@@ -0,0 +1,25 @@
+/* { dg-do run } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_0 "42" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_1 "43" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_2 "44" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_ALL "45" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV "46" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS "47" } */
+
+#include <omp.h>
+#include <stdlib.h>
+
+int
+main (int argc, char *const *argv)
+{
+ if (omp_get_max_teams () != 47)
+ abort ();
+
+ int num_devices = omp_get_num_devices () > 3 ? 3 : omp_get_num_devices ();
+ for (int i=0; i < num_devices; i++)
+ #pragma omp target device (i)
+ if (omp_get_max_teams () != 42 + i)
+ abort ();
+
+ return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/icv-6.c b/libgomp/testsuite/libgomp.c-c++-common/icv-6.c
new file mode 100644
index 0000000..05f07c7
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/icv-6.c
@@ -0,0 +1,44 @@
+/* { dg-do run } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_ALL "42" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV "43" } */
+/* { dg-set-target-env-var OMP_SCHEDULE_ALL "guided,4" } */
+/* { dg-set-target-env-var OMP_DYNAMIC_ALL "true" } */
+/* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_ALL "44" } */
+/* { dg-set-target-env-var OMP_THREAD_LIMIT_ALL "45" } */
+/* { dg-set-target-env-var OMP_NUM_THREADS_ALL "46,3,2" } */
+/* { dg-set-target-env-var OMP_MAX_ACTIVE_LEVELS_ALL "47" } */
+/* { dg-set-target-env-var OMP_PROC_BIND_ALL "spread" } */
+/* { dg-set-target-env-var OMP_WAIT_POLICY_ALL "active" } */
+
+/* This tests the hierarchical usage of ICVs on the device, i.e. if
+ OMP_NUM_TEAMS_DEV_<device_num> is not configured, then the value of
+ OMP_NUM_TEAMS_DEV should be used. And if */
+
+#include <omp.h>
+#include <stdlib.h>
+
+int
+main (int argc, char *const *argv)
+{
+ enum omp_sched_t kind;
+ int chunk_size;
+ omp_get_schedule(&kind, &chunk_size);
+
+ if (omp_get_max_teams () != 42
+ || !omp_get_dynamic ()
+ || kind != 3 || chunk_size != 4
+ || omp_get_teams_thread_limit () != 44
+ || omp_get_thread_limit () != 45
+ || omp_get_max_threads () != 46
+ || omp_get_proc_bind () != omp_proc_bind_spread
+ || omp_get_max_active_levels () != 47)
+ abort ();
+
+ int num_devices = omp_get_num_devices () > 3 ? 3 : omp_get_num_devices ();
+ for (int i=0; i < num_devices; i++)
+ #pragma omp target device (i)
+ if (omp_get_max_teams () != 43)
+ abort ();
+
+ return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/icv-7.c b/libgomp/testsuite/libgomp.c-c++-common/icv-7.c
new file mode 100644
index 0000000..67081dc
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/icv-7.c
@@ -0,0 +1,26 @@
+/* { dg-do run } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_ALL "42" } */
+
+/* This tests the hierarchical usage of ICVs on the host and on devices, i.e. if
+ OMP_NUM_TEAMS_DEV_<device_num>, OMP_NUM_TEAMS_DEV, and
+ OMP_NUM_TEAMS are not configured, then the value of
+ OMP_NUM_TEAMS_ALL should be used for the host as well as for the
+ devices. */
+
+#include <omp.h>
+#include <stdlib.h>
+
+int
+main (int argc, char *const *argv)
+{
+ if (omp_get_max_teams () != 42)
+ abort ();
+
+ int num_devices = omp_get_num_devices () > 3 ? 3 : omp_get_num_devices ();
+ for (int i=0; i < num_devices; i++)
+ #pragma omp target device (i)
+ if (omp_get_max_teams () != 42)
+ abort ();
+
+ return 0;
+}
diff --git a/libgomp/testsuite/libgomp.c-c++-common/icv-8.c b/libgomp/testsuite/libgomp.c-c++-common/icv-8.c
new file mode 100644
index 0000000..adaff5a
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/icv-8.c
@@ -0,0 +1,22 @@
+/* { dg-do run } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_1234567890 "42" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_ "43" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_01 "44" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_a "45" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_12345678901 "46" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_-1 "47" } */
+
+#include <omp.h>
+#include <stdlib.h>
+
+int
+main (int argc, char *const *argv)
+{
+ return 0;
+}
+
+/* { dg-output ".*Invalid device number in OMP_NUM_TEAMS_DEV_=43.*" } */
+/* { dg-output ".*Invalid device number in OMP_NUM_TEAMS_DEV_01=44 (leading zero).*" } */
+/* { dg-output ".*Invalid device number in OMP_NUM_TEAMS_DEV_a=45.*" } */
+/* { dg-output ".*Invalid device number in OMP_NUM_TEAMS_DEV_12345678901=46 (too long).*" } */
+/* { dg-output ".*Non-negative device number expected in OMP_NUM_TEAMS_DEV_-1=47.*" } */
diff --git a/libgomp/testsuite/libgomp.c-c++-common/omp-display-env-1.c b/libgomp/testsuite/libgomp.c-c++-common/omp-display-env-1.c
new file mode 100644
index 0000000..c586d3b
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/omp-display-env-1.c
@@ -0,0 +1,119 @@
+/* { dg-do run } */
+/* { dg-set-target-env-var OMP_THREAD_LIMIT_DEV_24 "42" } */
+/* { dg-set-target-env-var OMP_THREAD_LIMIT_ALL "43" } */
+/* { dg-set-target-env-var OMP_THREAD_LIMIT_DEV "44" } */
+/* { dg-set-target-env-var OMP_THREAD_LIMIT "45" } */
+/* { dg-set-target-env-var OMP_DEFAULT_DEVICE "42" } */
+/* { dg-set-target-env-var OMP_SCHEDULE_DEV_24 "guided,4" } */
+/* { dg-set-target-env-var OMP_SCHEDULE_ALL "dynamic" } */
+/* { dg-set-target-env-var OMP_SCHEDULE_DEV "guided,1" } */
+/* { dg-set-target-env-var OMP_SCHEDULE "guided,2" } */
+/* { dg-set-target-env-var OMP_DYNAMIC_DEV_24 "true" } */
+
+/* { dg-set-target-env-var OMP_DYNAMIC_ALL "true" } */
+/* { dg-set-target-env-var OMP_DYNAMIC_DEV "true" } */
+/* { dg-set-target-env-var OMP_DYNAMIC "true" } */
+/* { dg-set-target-env-var OMP_NUM_THREADS "4,3,2" } */
+/* { dg-set-target-env-var OMP_NUM_THREADS_ALL "45,46,47" } */
+/* { dg-set-target-env-var OMP_NUM_THREADS_DEV "42,43,44" } */
+/* { dg-set-target-env-var OMP_NUM_THREADS_DEV_24 "14,13,12" } */
+/* { dg-set-target-env-var OMP_MAX_ACTIVE_LEVELS "42" } */
+/* { dg-set-target-env-var OMP_MAX_ACTIVE_LEVELS_ALL "43" } */
+/* { dg-set-target-env-var OMP_MAX_ACTIVE_LEVELS_DEV "44" } */
+
+/* { dg-set-target-env-var OMP_MAX_ACTIVE_LEVELS_DEV_24 "45" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS "42" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_ALL "43" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV "44" } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS_DEV_24 "45" } */
+/* { dg-set-target-env-var OMP_PROC_BIND "spread" } */
+/* { dg-set-target-env-var OMP_PROC_BIND_ALL "close" } */
+/* { dg-set-target-env-var OMP_PROC_BIND_DEV "spread,spread" } */
+/* { dg-set-target-env-var OMP_PROC_BIND_DEV_24 "spread,close" } */
+/* { dg-set-target-env-var OMP_STACKSIZE "42" } */
+
+/* { dg-set-target-env-var OMP_STACKSIZE_ALL "42 M" } */
+/* { dg-set-target-env-var OMP_STACKSIZE_DEV "43 k" } */
+/* { dg-set-target-env-var OMP_STACKSIZE_DEV_24 "44" } */
+/* { dg-set-target-env-var OMP_WAIT_POLICY "active" } */
+/* { dg-set-target-env-var OMP_WAIT_POLICY_ALL "ACTIVE" } */
+/* { dg-set-target-env-var OMP_WAIT_POLICY_DEV "passive" } */
+/* { dg-set-target-env-var OMP_WAIT_POLICY_DEV_24 "PASSIVE" } */
+/* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT "42" } */
+/* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_ALL "43" } */
+/* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_DEV "44" } */
+
+/* { dg-set-target-env-var OMP_TEAMS_THREAD_LIMIT_DEV_24 "45" } */
+/* { dg-set-target-env-var OMP_CANCELLATION "true" } */
+/* { dg-set-target-env-var OMP_DISPLAY_AFFINITY "true" } */
+/* { dg-set-target-env-var OMP_TARGET_OFFLOAD "mandatory" } */
+/* { dg-set-target-env-var OMP_MAX_TASK_PRIORITY "20" } */
+/* { dg-set-target-env-var OMP_ALLOCATOR "omp_const_mem_alloc" } */
+/* { dg-set-target-env-var OMP_NESTED "false" } */
+
+#include <omp.h>
+#include <stdlib.h>
+
+int
+main (int argc, char *const *argv)
+{
+ omp_display_env (1);
+ return 0;
+}
+
+/* { dg-output ".*\\\[host] OMP_DYNAMIC = 'TRUE'.*" } */
+/* { dg-output ".*\\\[all] OMP_DYNAMIC = 'TRUE'.*" } */
+/* { dg-output ".*\\\[device] OMP_DYNAMIC = 'TRUE'.*" } */
+/* { dg-output ".*\\\[24\] OMP_DYNAMIC = 'TRUE'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_NUM_THREADS = '4,3,2'.*" } */
+/* { dg-output ".*\\\[all\] OMP_NUM_THREADS = '45,46,47'.*" } */
+/* { dg-output ".*\\\[device\] OMP_NUM_THREADS = '42,43,44'.*" } */
+/* { dg-output ".*\\\[24\] OMP_NUM_THREADS = '14,13,12'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_SCHEDULE = 'GUIDED,2'.*" } */
+/* { dg-output ".*\\\[all\] OMP_SCHEDULE = 'DYNAMIC'.*" } */
+/* { dg-output ".*\\\[device\] OMP_SCHEDULE = 'GUIDED'.*" } */
+/* { dg-output ".*\\\[24\] OMP_SCHEDULE = 'GUIDED,4'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_PROC_BIND = 'SPREAD'.*" } */
+/* { dg-output ".*\\\[all\] OMP_PROC_BIND = 'CLOSE'.*" } */
+/* { dg-output ".*\\\[device\] OMP_PROC_BIND = 'SPREAD,SPREAD'.*" } */
+/* { dg-output ".*\\\[24\] OMP_PROC_BIND = 'SPREAD,CLOSE'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_STACKSIZE = '43008'.*" } */
+/* { dg-output ".*\\\[all\] OMP_STACKSIZE = '44040192'.*" } */
+/* { dg-output ".*\\\[device\] OMP_STACKSIZE = '44032'.*" } */
+/* { dg-output ".*\\\[24\] OMP_STACKSIZE = '45056'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_WAIT_POLICY = 'ACTIVE'.*" } */
+/* { dg-output ".*\\\[all\] OMP_WAIT_POLICY = 'ACTIVE'.*" } */
+/* { dg-output ".*\\\[device\] OMP_WAIT_POLICY = 'PASSIVE'.*" } */
+/* { dg-output ".*\\\[24\] OMP_WAIT_POLICY = 'PASSIVE'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_THREAD_LIMIT = '45'.*" } */
+/* { dg-output ".*\\\[all\] OMP_THREAD_LIMIT = '43'.*" } */
+/* { dg-output ".*\\\[device\] OMP_THREAD_LIMIT = '44'.*" } */
+/* { dg-output ".*\\\[24\] OMP_THREAD_LIMIT = '42'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_MAX_ACTIVE_LEVELS = '42'.*" } */
+/* { dg-output ".*\\\[all\] OMP_MAX_ACTIVE_LEVELS = '43'.*" } */
+/* { dg-output ".*\\\[device\] OMP_MAX_ACTIVE_LEVELS = '44'.*" } */
+/* { dg-output ".*\\\[24\] OMP_MAX_ACTIVE_LEVELS = '45'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_NUM_TEAMS = '42'.*" } */
+/* { dg-output ".*\\\[all\] OMP_NUM_TEAMS = '43'.*" } */
+/* { dg-output ".*\\\[device\] OMP_NUM_TEAMS = '44'.*" } */
+/* { dg-output ".*\\\[24\] OMP_NUM_TEAMS = '45'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_TEAMS_THREAD_LIMIT = '42'.*" } */
+/* { dg-output ".*\\\[all\] OMP_TEAMS_THREAD_LIMIT = '43'.*" } */
+/* { dg-output ".*\\\[device\] OMP_TEAMS_THREAD_LIMIT = '44'.*" } */
+/* { dg-output ".*\\\[24\] OMP_TEAMS_THREAD_LIMIT = '45'.*" } */
+
+/* { dg-output ".*\\\[host] OMP_CANCELLATION = 'TRUE'.*" } */
+/* { dg-output ".*\\\[host] OMP_DEFAULT_DEVICE = '42'.*" } */
+/* { dg-output ".*\\\[host] OMP_MAX_TASK_PRIORITY = '20'.*" } */
+/* { dg-output ".*\\\[host] OMP_DISPLAY_AFFINITY = 'TRUE'.*" } */
+/* { dg-output ".*\\\[host] OMP_ALLOCATOR = 'omp_const_mem_alloc'.*" } */
+/* { dg-output ".*\\\[host] OMP_TARGET_OFFLOAD = 'MANDATORY'.*" } */
diff --git a/libgomp/testsuite/libgomp.c-c++-common/omp-display-env-2.c b/libgomp/testsuite/libgomp.c-c++-common/omp-display-env-2.c
new file mode 100644
index 0000000..660da4a
--- /dev/null
+++ b/libgomp/testsuite/libgomp.c-c++-common/omp-display-env-2.c
@@ -0,0 +1,22 @@
+/* { dg-do run } */
+/* { dg-set-target-env-var OMP_NUM_TEAMS "42" } */
+
+/* This test checks if omp_display_env outputs the initial ICV values although
+ the value was updated. */
+
+#include <omp.h>
+#include <stdlib.h>
+
+int
+main (int argc, char *const *argv)
+{
+ omp_display_env (1);
+ omp_set_num_teams (24);
+ if (omp_get_max_teams () != 24)
+ abort ();
+ omp_display_env (1);
+
+ return 0;
+}
+
+/* { dg-output ".*\\\[host] OMP_NUM_TEAMS = '42'.*\\\[host] OMP_NUM_TEAMS = '42'" } */
next prev parent reply other threads:[~2022-08-02 7:52 UTC|newest]
Thread overview: 20+ messages / expand[flat|nested] mbox.gz Atom feed top
2022-01-18 15:22 Marcel Vollweiler
2022-01-18 16:10 ` [PATCH] " Marcel Vollweiler
2022-05-04 15:12 ` Jakub Jelinek
2022-05-04 15:52 ` Tobias Burnus
2022-06-10 13:59 ` Marcel Vollweiler
2022-06-30 11:40 ` Jakub Jelinek
2022-06-30 13:21 ` Jakub Jelinek
2022-06-30 17:13 ` Jakub Jelinek
2022-07-04 15:14 ` Jakub Jelinek
2022-07-25 13:38 ` Marcel Vollweiler
2022-08-02 7:52 ` Marcel Vollweiler [this message]
2022-08-22 15:35 ` Jakub Jelinek
2022-08-31 10:56 ` Marcel Vollweiler
2022-09-06 11:51 ` Jakub Jelinek
2022-09-09 20:50 ` Rainer Orth
2022-09-09 22:08 ` Jakub Jelinek
2022-09-09 22:13 ` Iain Sandoe
2022-09-09 22:17 ` Jakub Jelinek
2022-09-10 13:17 ` Iain Sandoe
2022-09-09 11:50 ` [committed] libgomp: Fix up OMP_PROC_BIND handling [PR106894] Jakub Jelinek
Reply instructions:
You may reply publicly to this message via plain-text email
using any one of the following methods:
* Save the following mbox file, import it into your mail client,
and reply-to-all from there: mbox
Avoid top-posting and favor interleaved quoting:
https://en.wikipedia.org/wiki/Posting_style#Interleaved_style
* Reply using the --to, --cc, and --in-reply-to
switches of git-send-email(1):
git send-email \
--in-reply-to=c33653d4-20e7-96c3-4b9f-61edf2628682@codesourcery.com \
--to=marcel@codesourcery.com \
--cc=gcc-patches@gcc.gnu.org \
--cc=jakub@redhat.com \
/path/to/YOUR_REPLY
https://kernel.org/pub/software/scm/git/docs/git-send-email.html
* If your mail client supports setting the In-Reply-To header
via mailto: links, try the mailto: link
Be sure your reply has a Subject: header at the top and a blank line
before the message body.
This is a public inbox, see mirroring instructions
for how to clone and mirror all data and code used for this inbox;
as well as URLs for read-only IMAP folder(s) and NNTP newsgroup(s).