search

kokkos / kokkos-kernels

Kokkos C++ Performance Portability Programming Ecosystem: Math Kernels - Provides BLAS, Sparse BLAS and Graph Kernels
Other
303 stars 96 forks source link

Work around kk_get_free_total_memory not being available for all backends. #1062

Open masterleinad opened 3 years ago

masterleinad commented 3 years ago

Some backends, e.g. SYCL don't allow querying the amount of free device space. Nevertheless, KokkosSPGEMM::compute_num_pool_chunks, https://github.com/kokkos/kokkos-kernels/blob/0a1740890e84a145d6567768a0102f90a5c26f16/src/sparse/impl/KokkosSparse_spgemm_impl.hpp#L795-L816 and https://github.com/kokkos/kokkos-kernels/blob/0a1740890e84a145d6567768a0102f90a5c26f16/src/sparse/impl/KokkosSparse_gauss_seidel_impl.hpp#L1029-L1045 use it. Thus, using these classes/functions doesn't work for all backends.

masterleinad commented 3 years ago

This appears also in https://github.com/trilinos/Trilinos/pull/9266.

srajama1 commented 3 years ago

Is there a non-SYCL (native) way to query this in the hardware? This is such a basic functionality, I am hoping there is some low level call.

masterleinad commented 3 years ago

I am not aware of any mechanism (and also not in OpenCL).

brian-kelley commented 3 years ago

The rewritten SpGEMM implementation won't need to use memory pools, but as a temporary workaround I'm seeing what happens if free memory is hardcoded to something reasonable like 256 MB.

masterleinad commented 2 years ago

sycl::ext::oneapi::level_zero::info::device::free_memory should be available with the Level Zero backend, see https://github.com/intel/llvm/blob/sycl/sycl/doc/extensions/LevelZeroBackend/LevelZeroBackend.md#51-device-information-descriptors.

masterleinad commented 1 year ago

This is implemented in the compiler in https://github.com/intel/llvm/pull/6604 (using a very similar implementation).