[Bug testsuite/81690] libgomp.c/{target-32,thread-limit-2}.c fail for nvptx: missing usleep

[burnus at gcc dot gnu.org]

https://gcc.gnu.org/bugzilla/show_bug.cgi?id=81690

Tobias Burnus <burnus at gcc dot gnu.org> changed:

           What    |Removed                     |Added
----------------------------------------------------------------------------
                 CC|                            |burnus at gcc dot gnu.org

--- Comment #4 from Tobias Burnus <burnus at gcc dot gnu.org> ---
Regarding the code in comment 1 (which additionally lacks a "void" in the
decl):
The omp_is_initial_device() is only resolved at run time - hence, I think the
linker still wants to see "usleep".


I was wondering whether one shouldn't clone the files and use:

/* { dg-do run { target { { ! openacc_nvidia_accel_selected } && { !
openacc_amdgcn_accel_selected } } } } */

And the clone without "!" but with "||".


Other remarks:

* libgomp.c/target-32.c fails at run time with:
    libgomp: GOMP_OFFLOAD_async_run unimplemented
as soon as one got passed the usleep issue (e.g. by replacing it by a no-op)

* libgomp.c/thread-limit-2.c compiles - but it doesn't do anything on the GPU
due to:
  #pragma omp target if (0)


Regarding the do_sleep replacement:

* One needs to use additionally __attribute__((always_inline)) otherwise, it
fails with:
  function 'do_sleep' has been referenced in offloaded code but hasn't been
marked to be included in the offloaded code

* If one tries it with:
   #pragma omp declare target
that fails to compile with


libgomp.c/thread-limit-2.c compiles - but it doesn't do anything on the GPU due
to:
  #pragma omp target if (0)
many errors of the kind:
  .omp_data_kinds.24 / .omp_data_sizes.26 has been referenced in offloaded code
but hasn't been marked to be included in the offloaded code