cuDF numba cuda 12 updates - #13337

Ideally there's a better way of doing this rather than relying on a relative path.

A relative path is appropriate but use os.path.join and not string concatenation. You could also import cudf or a neighboring module and compute the path relative to that?

Runtime imports can be expensive but importing cuda must be avoided as a side effect of importing the module in which this function resides. Another option could be passing the cuda module as an optional argument here.

A relative path is appropriate but use os.path.join and not string concatenation. You could also import cudf or a neighboring module and compute the path relative to that?

I’m confused why we follow a different logical path for compute capabilities ending in “a.” They’re not fundamentally different, just differently named.

IIRC, this logic accounts for a multitude of different build configurations for both CI and local builds from source. I am happy to follow up on this but given that this function is only being moved I'm hesitant to anything that could perturb it as part of this PR.

Just wanting to be precise in our comments here. Do we specifically mean the runtime rather than the “toolkit”?

The toolkit component of concern is NVVM. We only use the runtime to check what version NVVM is (assuming if we find the runtime, that it's the same version as NVVM) because NVVM provides no way to check what version it is. I find it less confusing to refer to "the toolkit" as it's the collection of components including the runtime and NVVM, amongst other things.

Do we need to expand this to include 12.1? Users could run CUDA 12 pip wheels on 12.1.

Should we try to fall back to parsing the line “Cuda compilation tools, release 11.6” if this mapping fails? I know it might be less safe than the current solution but I wish we could use that to eliminate the need for a specific mapping that must be updated with CUDA releases.

I agree with this sentiment. Similarly to above however I think it's better to tackle this particular update separately as this function is a move. That said the change should be simple and I can commit to having it in for this release.

Use a descriptive name.

Does ptxcompiler use a subprocess currently? I am surprised by this because this tends to be costly. If possible, we should avoid launching a subprocess at cudf import time.

Is there a way to use any other tool (cuda-python, rmm, ...) to know this version information? I assume the goal of using a subprocess is to avoid importing numba.cuda in this process...

Yes, ptxcompiler uses a subprocess. This command is vendored directly from it here. While doing all of our setup before importing numba.cuda is one goal, another requirement is that the process for getting the versions doesn't cuinit, otherwise it can interfere with the way dask initializes its networking. More generally we need to avoid cuinit fully during import.

As an aside it would be nice if we had some kind of small independent package we could list as a dependency that served as a source of truth for the driver and runtime versions, which we could also configure to avoid forking a subprocess, calling cuinit, etc. This would avoid a bunch of reinventing the wheel across cuda-python, numba, rmm, etc....

I believe cuda-python is supposed to be just this, but in practice we’ve seen various issues (not providing the right runtime version, among others). I am not sure but recent cuda-python versions may be better at this?

I tried putting a

from cuda import cuda, cudart
cuda.cuDriverGetVersion()
cudart.cudaRuntimeGetVersion()

At the top of cudf's __init__.py and ran test_no_cuinit.py, which passed. So I suppose it's a clean approach from the cuInit perspective. Happy to refactor as such here and see if things work.

RMM does seem to list some kind of "limitation" here and falls back to numba, but I'm not sure what it means exactly
https://github.com/rapidsai/rmm/blob/branch-23.06/python/rmm/_cuda/gpu.py#L81-L85

Think this is referring to issue ( NVIDIA/cuda-python#16 )

Namely the version cuda-python returns is hard-coded at build time as opposed to querying it at run time. So the version it returns may not reflect what the user has installed on their system (and will instead reflect how the binary was built). The issue explains why this is happening currently (though it still presents an issue with intended usage)

There's more details in the threads in PR ( rapidsai/rmm#946 ) as well

To close the loop on this I decided in 5cb0ce6 to just vendor the three functions to _ptxcompiler.py.

To summarize the way it works now, the user either has ptxcompiler or doesn't. If they do, use it to get the versions (this launches a subprocess). If they don't, use the vendored functions to do so (this also launches the same subprocess using the same command).

We are constrained to use a subprocess for three reasons:

1. We can not call cuinit on cudf's import without interfering with dask.
2. We can not directly use numba.cuda to obtain the versions because we need to finish configuring numba before numba.cuda is imported.
3. We can not utilize cuda.cuda due to constraint described above

Even with the approach above we need to leave in place a mechanism to disable the spawning of a subprocess in HPC environments where it is not safe to do so. In this case the user may provide an environment variable which disables the check and two more environment variables specifying the driver and runtime versions manually, and these are parsed and returned instead. These environment variables are the same ones, between ptxcompiler and our vendored version.

This needs to use os.path.join.

Can we name this a proper name? NUMBA_CHECK_VERSION_CMD or similar.