hipBLAS

[pvelesko]

Does anyone have any sort of hipBLAS built for CHIP-SPV even with limited support? @Kerilk
Needed by libCEED

[pengtu]

We have started making a port of hipBLAS over oneMKL at: https://github.com/Sarbojit2019/hipBLAS. It is still in the very early stage but we will get there. @Sarbojit2019

[pengtu]

@pvelesko @Sarbojit2019 @Kerilk @pjaaskel: In the current prototype, each time hipblasCreate and hipblasSetStream is called, we create new SYCL context, device and queue to the underlying HIP stream's Level Zero native handles:

https://github.com/Sarbojit2019/hipBLAS/blob/develop/library/src/oneApi_detail/deps/sycl_wrapper.cpp#L74

In addition to the waste of duplicated contexts and queues, it is also a problem that multiple SYCL contexts and queues may not synchronize properly for multiple hipblas calls even with the same hipblasHandle.

One way to fix this is to maintain a per task map of task SYCL context, device, and per HIP stream map of SYCL queue in the CHIP-SPV task. The hipblasCreate and hipblasSetStream can call a function to get the corresponding SYCL queue for each HIP stream. The map may need to be reference counted so it will release the SYCL queue when hipblasDestroy() is called for each hipblasHandle and the refence count reaches zero, or release them during the task exit cleanup.

Is this feasible and any alternative solutions?

[Sarbojit2019]

@pengtu,
hipGetBackendNativeHandles() CHIP-SPV API which will return native handles for a queue/stream. Runtime does not create multiple context if same queue is queried twice. syclPlatformCreate() is a sycl call, CHIP-SPV does not know about it hence I can't see why CHIP-SPV needs to maintain any map here, having said that I don't think application will call hipBlasCreate multiple time for same device. See : https://docs.nvidia.com/cuda/cublas/index.html#cublas-context.

Coming back to our case we can maintain map [hipStream_t:syclblasHandle] in sycl_wrapper.cpp file and return value from the map in case of already used stream is getting set as current stream again. I am failing to understand why do we need to maintain ref count etc. I will talk to you offline to understand your view better.

[pengtu]

On the cublas link, it specifically states that an application can create multiple handles to achieve parallelism. Quoted below: This approach allows the user to explicitly control the library setup when using multiple host threads and multiple GPUs. For example, the application can use cudaSetDevice() to associate different devices with different host threads and in each of those host threads it can initialize a unique handle to the cuBLAS library context, which will use the particular device associated with that host thread. Then, the cuBLAS library function calls made with different handle will automatically dispatch the computation to different devices.

…

On Sat, Dec 3, 2022 at 5:59 AM Sarbojit2019 ***@***.***> wrote: @pengtu <https://github.com/pengtu>, hipGetBackendNativeHandles() CHIP-SPV API which will return native handles for a queue/stream. Runtime does not create multiple context if same queue is queried twice. syclPlatformCreate() is a sycl call, CHIP-SPV does not know about it hence I can't see why CHIP-SPV needs to maintain any map here, having said that I don't think application will call hipBlasCreate multiple time for same device. See : https://docs.nvidia.com/cuda/cublas/index.html#cublas-context. Coming back to our case we can maintain map [hipStream_t:syclblasHandle] in sycl_wrapper.cpp file and return value from the map in case of already used stream is getting set as current stream again. I am failing to understand why do we need to maintain ref count etc. I will talk to you offline to understand your view better. — Reply to this email directly, view it on GitHub <#273 (reply in thread)>, or unsubscribe <https://github.com/notifications/unsubscribe-auth/AATZITC3G53SD333263G4GDWLNG3NANCNFSM6AAAAAASRZCMDQ> . You are receiving this because you were mentioned.Message ID: ***@***.***>

[pengtu]

We also need to consider the scenario when an application calls multiple libraries, for example, cublas and cufft. The cufft handle must share the same SYCL context of cublas if they access the same memory.

…

On Sun, Dec 4, 2022 at 6:32 AM Peng Tu ***@***.***> wrote: On the cublas link, it specifically states that an application can create multiple handles to achieve parallelism. Quoted below: This approach allows the user to explicitly control the library setup when using multiple host threads and multiple GPUs. For example, the application can use cudaSetDevice() to associate different devices with different host threads and in each of those host threads it can initialize a unique handle to the cuBLAS library context, which will use the particular device associated with that host thread. Then, the cuBLAS library function calls made with different handle will automatically dispatch the computation to different devices. On Sat, Dec 3, 2022 at 5:59 AM Sarbojit2019 ***@***.***> wrote: > @pengtu <https://github.com/pengtu>, > hipGetBackendNativeHandles() CHIP-SPV API which will return native > handles for a queue/stream. Runtime does not create multiple context if > same queue is queried twice. syclPlatformCreate() is a sycl call, CHIP-SPV > does not know about it hence I can't see why CHIP-SPV needs to maintain any > map here, having said that I don't think application will call > hipBlasCreate multiple time for same device. See : > https://docs.nvidia.com/cuda/cublas/index.html#cublas-context. > > Coming back to our case we can maintain map [hipStream_t:syclblasHandle] > in sycl_wrapper.cpp file and return value from the map in case of already > used stream is getting set as current stream again. I am failing to > understand why do we need to maintain ref count etc. I will talk to you > offline to understand your view better. > > — > Reply to this email directly, view it on GitHub > <#273 (reply in thread)>, > or unsubscribe > <https://github.com/notifications/unsubscribe-auth/AATZITC3G53SD333263G4GDWLNG3NANCNFSM6AAAAAASRZCMDQ> > . > You are receiving this because you were mentioned.Message ID: > ***@***.***> >

[pvelesko]

@Sarbojit2019 libCEED needs the following
Can you give an estimate on when these could be implemented?

hipblasCreate
hipblasDasum
hipblasDestroy
hipblasDnrm2
hipblasIdamax

[Sarbojit2019]

@pvelesko,
All of above mentioned APIs are implemented here https://github.com/Sarbojit2019/hipBLAS. I will send you separate mail with steps to build and how to use part.

Note: IIRC, libCEED is a C-library, due to CHIP-SPV 'C' incompatibility I could not verify it with C sample.