FileHandle::read_async()/FileHandle::write_async() CUstream and StreamFuture

[Jacobfaib]

TL;DR: Is it possible to expose a fire-and-forget-like version of the async read/write functions if a CUstream functions are given?

---

FileHandle::read_async() (and write_async()) state that these functions return a StreamFuture which

must be kept alive until all data has been read to disk

Ostensibly this is so that stream synchronization can be RAII-ified (StreamFuture indeed synchronizes the stream in its dtor).

But, if I pass a non-null CUstream then why must I keep the StreamFuture alive? If I have given kvikio a stream, then surely kvikio may assume that I will perform other things on the same stream.

For reference, I am thinking of the following possible workflow:

void *read_gpu_data(std::size_t nbytes, CUstream stream)
{
  auto f = kvikio::FileHandle{"data.txt", "r"};

  void *dev_ptr = allocate_gpu_pointer(nbytes);
  f.read_async(dev_ptr, nbytes, 0, stream);
  return dev_ptr;
}

void do_work()
{
  std::size_t nbytes = 1234;

  auto *dev_ptr = read_gpu_data(nbytes, stream);
  my_kernel<<<..., stream>>>(dev_ptr, ...);
}

In this example, if I want to keep the GPU pipeline alive, I also need to ferry around the StreamFuture object...

The other overload taking pointers is also not ergonomic, because it requires that bytes_read_p is non-NULL. I know the values of all my parameters right now, and I don't care about how many bytes were read by the GPU task (because I'm passing the pointer on to the kernel).

[madsbk]

The other overload taking pointers is also not ergonomic, because it requires that bytes_read_p is non-NULL.

This is the reason why KvikIO introduce the StreamFuture object. The only async API cuFile provides takes the arguments as pointers. There is no way to provide the arguments by value :/

[Jacobfaib]

This is the reason why KvikIO introduce the StreamFuture object. The only async API cuFile provides takes the arguments as pointers. There is no way to provide the arguments by value :/

Hmmm, perhaps kvikio could provide a "default" pointer to use here in that case. Some kind of internal static pointer that kvikio registers appropriately with cuFile which can be used as a default if this pointer is NULL.

[madsbk]

But to handle errors, we would need to have a bytes_read_p per call. We could use cuStreamAddCallback to clean up bytes_read_p et al. but what about the FileHandle?
In your example, the clean up of FileHandle also needs to be delayed. Is this required?