Allow for more data input types and structures

[djhoese]

The Problem

Delayed data loading is becoming more and more necessary as datasets grow larger and have to be accessed from remote storage (S3, etc). Projects like dask (and the communities that use it) make it easy to do parallel data processing and have made the idea of chunked/tiled asynchronous data very popular and useful. The efforts by the dask and numpy communities have resulted in a more defined specification for array subclasses (https://numpy.org/doc/stable/reference/arrays.classes.html#numpy.class.__array_function__) that have benefited dask, xarray, cupy, zarr, and many other libraries.

The problem is VisPy can still only naively use contiguous in-memory numpy arrays (ok, fine, memory maps are technically not in-memory) that it sends to the GPU. VisPy could be a huge resource for these various libraries if we could allow more generic access to the data they represent. For example, a tiled image source being provided to an ImageVisual and pushing the data to the GPU when it is ready or when it is viewed. Or a large complicated dask computation being performed on a cluster and being passed to a VolumeVisual or ImageVisual or MeshVisual and having the pieces show up in the visualization once they are done computing. Or a CuPy computation (maybe even being managed by dask) that can run on the GPU and then immediately draw that data with VisPy; no copy to CPU (see #1985).

The Options

If any of this stuff is possible already, let's get some examples or at least add information to the sphinx docs about projects that are doing these things (@codypiersall's vismap).

Other thoughts I've had are some kind of "data source" API/wrapper classes that can be passed to the various VisPy .set_data methods and allow Visuals to take advantage of these different data sources. I'm not sure if these data sources would have to manage a texture or buffer or if the Visual could tell the data source what it needs or allows.

CC @jni @codypiersall

Related:

• Add CuPy support #1985
• https://github.com/codypiersall/vismap

[jakirkham]

Have you looked into __cuda_array_interface__? This is what we have been using to share objects between different libraries (RAPIDS, CuPy, Numba, PyTorch, etc.)

[djhoese]

No, but that would definitely help. Thanks for pointing it out. My main problem now is how to give that underlying GPU buffer/array to OpenGL.

[trxcllnt]

@djhoese you could use the CUDA <-> OpenGL interop from the CUDA Runtime API:

1. create GL buffers like normal, register with CUDA via cudaGraphicsGLRegisterBuffer()
2. map GL buffers for writing via cudaGraphicsMapResources()
3. use cudaGraphicsResourceGetMappedPointer() to get a CUDA device pointer for each mapped GL buffer
4. call cudaMemcpy(gl_ptr, cu_ptr)
5. use cudaGraphicsUnmapResources() to unmap GL buffers
6. ???
7. profit

edit: the above assumes the CUDA and GL contexts both belong to the same process

[djhoese]

Nice! Thanks for the info @trxcllnt. A couple questions and comments:

• Regarding your edit, I think this is a safe enough assumption for now. I assume you are thinking something like dask creating a ton of GPU workers from separate processes which would not be possible with this solution, right?
• In your item 4, would that be a device-to-device copy? All on the GPU that is. Does that actually copy the data? Is there any way to do this without a copy?
• Looking around it looks like cudaGraphicsGLRegisterImage would be a good alternative for your number 1 in the cases where we'd want to get OpenGL texture data (VisPy's ImageVisual for example).
• Any idea if there is an interop layer for Vulkan? @rossant might be interested in that. (Nevermind: found this https://github.com/NVIDIA/cuda-samples/blob/master/Samples/simpleVulkan/README.md)

I'd have to figure out how to do all the CUDA context connecting, but this doesn't seem that bad for a proof of concept at least. I don't have a use case that would actually benefit from this, but that's the point of a proof of concept, right? 😉

[rossant]

@djhoese yeah there are ways to do GL-CUDA and Vulkan-CUDA interop without unnecessary array copying!
With Vulkan you can even more simply use compute shaders, I think they're more limited than CUDA but they still give you a way to launch GPU kernels on GPU buffers and textures. You have full control on the interaction between graphics and compute, you can handle the synchronization manually if you want, etc.

Now you don't necessarily have to use that. For example, if you want to implement something like Google Maps with infinite zooming in an image, with images being fetched on the fly, you could perhaps use an existing image visual and reupload the texture dynamically.

[quasiben]

I'd have to figure out how to do all the CUDA context connecting, but this doesn't seem that bad for a proof of concept at least. I don't have a use case that would actually benefit from this, but that's the point of a proof of concept, right?

A use case here would be, as you said, doing some Dask work with GPUs then handing off to visualize with VisPy. This is being discussed amongst Napari folks: napari/napari#2243 where VisPy is leveraged

The workflow we've experimented with looks like the following:

1. load zarr data with dask
2. convert to GPU
3. do some image filtering
4. convert back to numpy
5. visualize with napari

Removing step 4 and handing Dask GPU arrays directly to Napari would be very cool!

[djhoese]

I've never used dask with GPU work. Does it require using a non-threaded scheduler? If so, then what @trxcllnt said about sharing the GL and CUDA contexts might be a problem. Unless dask/cuda have a public interface for saying "I know this data is spread across multiple contexts/devices/nodes, but bring it all to this context as efficiently as possible".

[jakirkham]

AIUI users are not doing this on Workers, but are doing this through the Client. Or for the non-Distributed case this would be happening in the main (application) thread. More generally they are calling .compute(). So I don't think we have to worry about that

[djhoese]

@jakirkham I read the discussion at https://giters.com/cupy/cupy/issues/5711 as best I could (internal references aren't available to external users and just show __) and noticed that Leo Fang ended it with saying CUDA Python would have some of the necessary interfaces available. Do you know of any updates on this? What version specifically? What interfaces specifically?

[jakirkham]

Interesting that appears to be a clone of the GH issue ( cupy/cupy#5711 ).

In any event, I think what Leo was saying (and Leo please feel free to correct me) is a future version of CUDA Python would have cudaGraphicsGLRegisterBuffer (and possibly other CUDA GL functions I'm forgetting).

There is not a specific version as those changes have not been released yet.

cc @leofang

[djhoese]

Interesting that appears to be a clone of the GH issue ( cupy/cupy#5711 ).

Damn it google and these shell websites.