On 12-12-18 10:27, Jakub Jelinek wrote:
> On Wed, Dec 12, 2018 at 09:25:06AM +0100, Tom de Vries wrote:
>> [openmp] Fix openmp 5.0 builtin function types
>>
>> Fix some openmp 5.0 builtin functions to match the type used in the
>> implementation of those functions.
>>
>> This fixes some libgomp testsuite failures for x86_64 with nvptx accelerator.
>>
>> Build on x86_64 with nvptx accelerator, tested libgomp.
>>
>> 2018-12-12 Tom de Vries <tdevr...@suse.de>
>>
>> * omp-builtins.def
>> (BUILT_IN_GOMP_LOOP_MAYBE_NONMONOTONIC_RUNTIME_START)
>> (BUILT_IN_GOMP_LOOP_ULL_MAYBE_NONMONOTONIC_RUNTIME_START)
>> (BUILT_IN_GOMP_PARALLEL_LOOP_MAYBE_NONMONOTONIC_RUNTIME): Fix
>> function type.
>
> Good catch, but it looks insufficient, the builtins without _maybe have the
> same bug (copy and paste from the preceding dynamic/guided calls that do
> have that extra argument - chunk_size), so if you've caught only these,
> it seems we have insufficient testsuite covergate in the target regions.
>
> Does the following testcase show also the remaining three bugs on unpatched
> nvptx offloading
It does, I get "Call has wrong number of parameters" errors for these 3
mismatches:
...
.extern .func (.param .u32 %value_out)
GOMP_loop_nonmonotonic_runtime_start (.param .u64 %in_ar0, .param .u64
%in_ar1, .param .u64 %in_ar2, .param .u64 %in_ar3, .param .u64 %in_ar4,
.param .u64 %in_ar5);
call
(%value_in),GOMP_loop_nonmonotonic_runtime_start,(%out_arg1,%out_arg2,%out_arg3,%out_arg4,%out_arg5);
.extern .func (.param .u32 %value_out)
GOMP_loop_ull_nonmonotonic_runtime_start (.param .u32 %in_ar0, .param
.u64 %in_ar1, .param .u64 %in_ar2, .param .u64 %in_ar3, .param .u64
%in_ar4, .param .u64 %in_ar5, .param .u64 %in_ar6);
call
(%value_in),GOMP_loop_ull_nonmonotonic_runtime_start,(%out_arg1,%out_arg2,%out_arg3,%out_arg4,%out_arg5,%out_arg6);
.extern .func GOMP_parallel_loop_nonmonotonic_runtime (.param .u64
%in_ar0, .param .u64 %in_ar1, .param .u32 %in_ar2, .param .u64 %in_ar3,
.param .u64 %in_ar4, .param .u64 %in_ar5, .param .u64 %in_ar6, .param
.u32 %in_ar7);
call
GOMP_parallel_loop_nonmonotonic_runtime,(%out_arg1,%out_arg2,%out_arg3,%out_arg4,%out_arg5,%out_arg6,%out_arg7);
...
and after rebuilding the test passes.
> (sorry, don't have nvptx offloading set up right now,
> should fix that soon)?
>
Np.
Thanks,
- Tom
> 2018-12-12 Tom de Vries <tdevr...@suse.de>
> Jakub Jelinek <ja...@redhat.com>
>
> * omp-builtins.def (BUILT_IN_GOMP_LOOP_NONMONOTONIC_RUNTIME_START,
> BUILT_IN_GOMP_LOOP_MAYBE_NONMONOTONIC_RUNTIME_START,
> BUILT_IN_GOMP_LOOP_ULL_NONMONOTONIC_RUNTIME_START,
> BUILT_IN_GOMP_LOOP_ULL_MAYBE_NONMONOTONIC_RUNTIME_START,
> BUILT_IN_GOMP_PARALLEL_LOOP_NONMONOTONIC_RUNTIME,
> BUILT_IN_GOMP_PARALLEL_LOOP_MAYBE_NONMONOTONIC_RUNTIME): Fix up
> function types - remove one argument.
>
> * testsuite/libgomp.c-c++-common/for-16.c: New test.
>
> --- gcc/omp-builtins.def.jj 2018-11-08 18:07:56.345070635 +0100
> +++ gcc/omp-builtins.def 2018-12-12 10:03:14.355354282 +0100
> @@ -126,11 +126,11 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_NON
> ATTR_NOTHROW_LEAF_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_NONMONOTONIC_RUNTIME_START,
> "GOMP_loop_nonmonotonic_runtime_start",
> - BT_FN_BOOL_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR,
> + BT_FN_BOOL_LONG_LONG_LONG_LONGPTR_LONGPTR,
> ATTR_NOTHROW_LEAF_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_MAYBE_NONMONOTONIC_RUNTIME_START,
> "GOMP_loop_maybe_nonmonotonic_runtime_start",
> - BT_FN_BOOL_LONG_LONG_LONG_LONG_LONGPTR_LONGPTR,
> + BT_FN_BOOL_LONG_LONG_LONG_LONGPTR_LONGPTR,
> ATTR_NOTHROW_LEAF_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ORDERED_STATIC_START,
> "GOMP_loop_ordered_static_start",
> @@ -234,11 +234,11 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL
> ATTR_NOTHROW_LEAF_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_NONMONOTONIC_RUNTIME_START,
> "GOMP_loop_ull_nonmonotonic_runtime_start",
> - BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULL_ULLPTR_ULLPTR,
> + BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULLPTR_ULLPTR,
> ATTR_NOTHROW_LEAF_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_MAYBE_NONMONOTONIC_RUNTIME_START,
> "GOMP_loop_ull_maybe_nonmonotonic_runtime_start",
> - BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULL_ULLPTR_ULLPTR,
> + BT_FN_BOOL_BOOL_ULL_ULL_ULL_ULLPTR_ULLPTR,
> ATTR_NOTHROW_LEAF_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_ULL_ORDERED_STATIC_START,
> "GOMP_loop_ull_ordered_static_start",
> @@ -349,11 +349,11 @@ DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL
> ATTR_NOTHROW_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL_LOOP_NONMONOTONIC_RUNTIME,
> "GOMP_parallel_loop_nonmonotonic_runtime",
> - BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_LONG_UINT,
> + BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_UINT,
> ATTR_NOTHROW_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_PARALLEL_LOOP_MAYBE_NONMONOTONIC_RUNTIME,
> "GOMP_parallel_loop_maybe_nonmonotonic_runtime",
> - BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_LONG_UINT,
> + BT_FN_VOID_OMPFN_PTR_UINT_LONG_LONG_LONG_UINT,
> ATTR_NOTHROW_LIST)
> DEF_GOMP_BUILTIN (BUILT_IN_GOMP_LOOP_END, "GOMP_loop_end",
> BT_FN_VOID, ATTR_NOTHROW_LEAF_LIST)
> --- libgomp/testsuite/libgomp.c-c++-common/for-16.c.jj 2018-12-12
> 10:19:50.640074341 +0100
> +++ libgomp/testsuite/libgomp.c-c++-common/for-16.c 2018-12-12
> 10:19:22.100540708 +0100
> @@ -0,0 +1,114 @@
> +extern
> +#ifdef __cplusplus
> +"C"
> +#endif
> +void abort (void);
> +
> +unsigned long long int k = 16;
> +#pragma omp declare target to (k)
> +
> +int
> +main ()
> +{
> + unsigned char a[144], b[144], c[144];
> + int l;
> + #pragma omp target map(from:a, b, c)
> + {
> + int i;
> + unsigned long long int j;
> + #pragma omp parallel for schedule (runtime)
> + for (i = 0; i < 16; i++)
> + a[i] = i;
> + #pragma omp parallel for schedule (monotonic: runtime)
> + for (i = 0; i < 16; i++)
> + a[i + 16] = i + 16;
> + #pragma omp parallel for schedule (nonmonotonic: runtime)
> + for (i = 0; i < 16; i++)
> + a[i + 32] = i + 32;
> + #pragma omp parallel
> + {
> + #pragma omp for schedule (runtime)
> + for (i = 0; i < 16; i++)
> + a[i + 48] = i + 48;
> + #pragma omp for schedule (monotonic: runtime)
> + for (i = 0; i < 16; i++)
> + a[i + 64] = i + 64;
> + #pragma omp for schedule (nonmonotonic: runtime)
> + for (i = 0; i < 16; i++)
> + a[i + 80] = i + 80;
> + #pragma omp for schedule (runtime)
> + for (j = 0; j < k; j++)
> + a[j + 96] = j + 96;
> + #pragma omp for schedule (monotonic: runtime)
> + for (j = 0; j < k; j++)
> + a[j + 112] = j + 112;
> + #pragma omp for schedule (nonmonotonic: runtime)
> + for (j = 0; j < k; j++)
> + a[j + 128] = j + 128;
> + }
> + #pragma omp parallel for schedule (dynamic)
> + for (i = 0; i < 16; i++)
> + b[i] = i;
> + #pragma omp parallel for schedule (monotonic: dynamic)
> + for (i = 0; i < 16; i++)
> + b[i + 16] = i + 16;
> + #pragma omp parallel for schedule (nonmonotonic: dynamic)
> + for (i = 0; i < 16; i++)
> + b[i + 32] = i + 32;
> + #pragma omp parallel
> + {
> + #pragma omp for schedule (dynamic)
> + for (i = 0; i < 16; i++)
> + b[i + 48] = i + 48;
> + #pragma omp for schedule (monotonic: dynamic)
> + for (i = 0; i < 16; i++)
> + b[i + 64] = i + 64;
> + #pragma omp for schedule (nonmonotonic: dynamic)
> + for (i = 0; i < 16; i++)
> + b[i + 80] = i + 80;
> + #pragma omp for schedule (dynamic)
> + for (j = 0; j < k; j++)
> + b[j + 96] = j + 96;
> + #pragma omp for schedule (monotonic: dynamic)
> + for (j = 0; j < k; j++)
> + b[j + 112] = j + 112;
> + #pragma omp for schedule (nonmonotonic: dynamic)
> + for (j = 0; j < k; j++)
> + b[j + 128] = j + 128;
> + }
> + #pragma omp parallel for schedule (guided)
> + for (i = 0; i < 16; i++)
> + c[i] = i;
> + #pragma omp parallel for schedule (monotonic: guided)
> + for (i = 0; i < 16; i++)
> + c[i + 16] = i + 16;
> + #pragma omp parallel for schedule (nonmonotonic: guided)
> + for (i = 0; i < 16; i++)
> + c[i + 32] = i + 32;
> + #pragma omp parallel
> + {
> + #pragma omp for schedule (guided)
> + for (i = 0; i < 16; i++)
> + c[i + 48] = i + 48;
> + #pragma omp for schedule (monotonic: guided)
> + for (i = 0; i < 16; i++)
> + c[i + 64] = i + 64;
> + #pragma omp for schedule (nonmonotonic: guided)
> + for (i = 0; i < 16; i++)
> + c[i + 80] = i + 80;
> + #pragma omp for schedule (guided)
> + for (j = 0; j < k; j++)
> + c[j + 96] = j + 96;
> + #pragma omp for schedule (monotonic: guided)
> + for (j = 0; j < k; j++)
> + c[j + 112] = j + 112;
> + #pragma omp for schedule (nonmonotonic: guided)
> + for (j = 0; j < k; j++)
> + c[j + 128] = j + 128;
> + }
> + }
> + for (l = 0; l < 144; ++l)
> + if (a[l] != l || b[l] != l || c[l] != l)
> + abort ();
> + return 0;
> +}
>
>
> Jakub
>
