Fix #7041: Add charseq registry to CUDA target - #7057

Unsure of what milestone to suggest - this should be a short PR to deal with, but I'm aware the queue for 0.54 is already insanely long.

@stuartarchibald Many thanks for the review - suggestion applied.

Buildfarm ID: numba_smoketest_cuda_yaml_71.

Buildfarm ID: numba_smoketest_cuda_yaml_71.

Failed, nvvm segfault for test numba.cuda.tests.cudapy.test_const_string.py::test_assign_const_string, there's no obvious pattern to the failures in the test matrix.

Failed, nvvm segfault for test numba.cuda.tests.cudapy.test_const_string.py::test_assign_const_string, there's no obvious pattern to the failures in the test matrix.

@stuartarchibald Is it possible that CUDA toolkits < 11.2 failed and CUDA toolkits >= 11.2 passed?

Failed, nvvm segfault for test numba.cuda.tests.cudapy.test_const_string.py::test_assign_const_string, there's no obvious pattern to the failures in the test matrix.

@stuartarchibald Is it possible that CUDA toolkits < 11.2 failed and CUDA toolkits >= 11.2 passed?

Output was:

Linux:
Passed: (py39, np1.20, cuda 11.2)
Failed: (py39, np1.20, cuda 10.0)
Failed: (py39, np1.20, cuda 9.2)

Passed: (py38, np1.17, cuda 11.2)
Failed: (py38, np1.17, cuda 10.0)
Passed: (py38, np1.17, cuda 9.2)

On windows all the combinations above failed except (py38, np1.17, cuda 11.2)

Thanks for the details - I'm struggling to replicate the failure on Windows with (py39, np1.20, cuda 11.2) - my environment is:

(numbapy39np120) C:\work\numbadev\numba [grm-issue-7041 ≡]> conda list
# packages in environment at C:\work\miniconda3\envs\numbapy39np120:
#
# Name                    Version                   Build  Channel
backcall                  0.2.0              pyhd3eb1b0_0
blas                      1.0                         mkl
ca-certificates           2021.7.5             haa95532_1
certifi                   2021.5.30        py39haa95532_0
cffi                      1.14.6           py39h2bbff1b_0
colorama                  0.4.4              pyhd3eb1b0_0
cudatoolkit               11.2.2               h933977f_8    conda-forge
decorator                 5.0.9              pyhd3eb1b0_0
intel-openmp              2021.3.0          haa95532_3372
ipython                   7.26.0           py39hd4e2768_0
ipython_genutils          0.2.0              pyhd3eb1b0_1
jedi                      0.18.0           py39haa95532_1
llvmdev                   11.1.0                        2    numba
llvmlite                  0.38.0.dev0+14.g09df344           dev_0    <develop>
matplotlib-inline         0.1.2              pyhd3eb1b0_2
mkl                       2021.3.0           haa95532_524
mkl-service               2.4.0            py39h2bbff1b_0
mkl_fft                   1.3.0            py39h277e83a_2
mkl_random                1.2.2            py39hf11a4ad_0
numba                     0.54.0.dev0+636.gb52c31e4d           dev_0    <develop>
numpy                     1.20.3           py39ha4e8547_0
numpy-base                1.20.3           py39hc2deb75_0
openssl                   1.1.1k               h2bbff1b_0
parso                     0.8.2              pyhd3eb1b0_0
pickleshare               0.7.5           pyhd3eb1b0_1003
pip                       21.2.2           py39haa95532_0
prompt-toolkit            3.0.17             pyh06a4308_0
pycparser                 2.20                       py_2
pygments                  2.9.0              pyhd3eb1b0_0
python                    3.9.6                h6244533_0
setuptools                52.0.0           py39haa95532_0
six                       1.16.0             pyhd3eb1b0_0
sqlite                    3.36.0               h2bbff1b_0
traitlets                 5.0.5              pyhd3eb1b0_0
tzdata                    2021a                h52ac0ba_0
vc                        14.2                 h21ff451_1
vs2015_runtime            14.27.29016          h5e58377_2
wcwidth                   0.2.5                      py_0
wheel                     0.36.2             pyhd3eb1b0_0
wincertstore              0.2              py39h2bbff1b_0

Do you see any discrepancies between my setup and the failing Windows + CUDA 11.2 setup? If not, do you have any more details about the failure? (e.g. a .dmp file, some details from the event log at the time of the crash, any traceback presented, etc.)?

My hypothesis was that this is an NVVM 3.4 vs NVVM 7.0 thing - i.e. that the failures were running into an NVVM 3.4 bug - because its a memory corruption, it could sometimes pass on older toolkit versions by luck, like the (py38, np1.17, cuda 9.2) instance on Linux. If that proved to be the case, my plan was to only support charseqs and anything that builds on it when using CUDA 11.2 onwards, and strongly recommend the use of 11.2 with Numba from this point onwards (given that also lineinfo / debugging support and other things are also falling behind with toolkits < 11.2).

@stuartarchibald Thanks for the legalization suggestion. I've gone with it, but made it raise instead of warn, since character sequences dangerous on NVVM 3.4 and never worked before anyway. I also created #7579 to record the fact that this still doesn't enable character sequences in record on CUDA.

I think this is ready for another look now.

@gmarkall No problem. I agree with the raise instead of warn, given it is a dangerous/unpredictable route to go down. Should a note be added to the docs to this effect i.e. that with CTK 11.2+ the CUDA target now has charseq support?

gpuci run tests

gpuci run tests

gpuci run tests

Buildfarm ID: numba_smoketest_cuda_yaml_103.