Skip to content

Update CtypesLinker::add_cu error message to include fp16 usage #8744

New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

Merged
sklam merged 14 commits into from
Mar 28, 2023
Merged

Update CtypesLinker::add_cu error message to include fp16 usage #8744

Show file tree
Hide file tree
Changes from 6 commits
Commits
Show all changes
14 commits
Select commit Hold shift + click to select a range
aec2c5d
Update error message to include f16 usage
testhound Feb 8, 2023
12288b8
Fix flake8 error
testhound Feb 8, 2023
369b016
Merge branch 'main' into testhound/cu_linking_err_msg
testhound Feb 13, 2023
7704080
Update error message to be more informative
testhound Feb 15, 2023
db210b6
Separate out float16 error exception from cu linking
testhound Feb 16, 2023
275243e
Run test case unconditionally, fix file not being found and override …
testhound Feb 16, 2023
5212ac6
Add new test to verify float16 exception
testhound Feb 28, 2023
467daf6
Merge branch 'main' into testhound/cu_linking_err_msg
testhound Feb 28, 2023
dad4421
Fix typo
testhound Feb 28, 2023
6587af4
Fix typo and simplify test case
testhound Feb 28, 2023
932046d
Skip test if minor version compatability used
testhound Feb 28, 2023
4e4ecd8
Merge branch 'main' into testhound/cu_linking_err_msg
testhound Mar 17, 2023
51f914d
Merge remote-tracking branch 'numba/main' into testhound/cu_linking_e…
gmarkall Mar 28, 2023
18dae59
CUDA test_linker.py: fix jitlink.cu path
gmarkall Mar 28, 2023
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
9 changes: 9 additions & 0 deletions numba/cuda/dispatcher.py
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -110,6 +110,15 @@ def __init__(self, py_func, argtypes, link=None, debug=False,
if (f'__numba_wrapper_{fn}' in lib.get_asm_str())]

if res:
if not config.CUDA_USE_NVIDIA_BINDING:
s = "https://numba.readthedocs.io/en/stable/cuda/bindings.html"
msg = ("Use of float16 requires the use of the NVIDIA CUDA "
"bindings and settng the "
Comment thread
gmarkall marked this conversation as resolved.
Resolved
"NUMBA_CUDA_USE_NVIDIA_BINDING environment variable to "
"1. Relevant documentation is available here:\n"
f"{s}")
Comment on lines +115 to +119

@gmarkall gmarkall Feb 17, 2023

Copy link
Copy Markdown
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Some simplification of this comment (and the typo fix suggested yesterday incorporated):

Suggested change
msg = ("Use of float16 requires the use of the NVIDIA CUDA "
"bindings and settng the "
"NUMBA_CUDA_USE_NVIDIA_BINDING environment variable to "
"1. Relevant documentation is available here:\n"
f"{s}")
msg = ("Use of float16 requires the NVIDIA CUDA bindings and "
"setting the NUMBA_CUDA_USE_NVIDIA_BINDING environment "
f"variable to 1. See {s}")

raise NotImplementedError(msg)

# Path to the source containing the foreign function

basedir = os.path.dirname(os.path.abspath(__file__))
Expand Down
18 changes: 13 additions & 5 deletions numba/cuda/tests/cudadrv/test_linker.py
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -3,12 +3,11 @@
import warnings
from numba.cuda.testing import unittest
from numba.cuda.testing import (skip_on_cudasim, skip_unless_cuda_python,
skip_if_cuda_includes_missing,
skip_with_cuda_python)
skip_if_cuda_includes_missing)
from numba.cuda.testing import CUDATestCase
from numba.cuda.cudadrv.driver import Linker, LinkerError, NvrtcError
from numba.cuda import require_context
from numba.tests.support import ignore_internal_warnings
from numba.tests.support import TestCase, ignore_internal_warnings
from numba import cuda, void, float64, int64, int32, typeof

CONST1D = np.arange(10, dtype=np.float64)
Expand Down Expand Up @@ -92,6 +91,7 @@ def simple_lmem(A, B, dty):

@skip_on_cudasim('Linking unsupported in the simulator')
class TestLinker(CUDATestCase):
_NUMBA_NVIDIA_BINDING_0_ENV = {'NUMBA_CUDA_USE_NVIDIA_BINDING': '0'}

@require_context
def test_linker_basic(self):
Expand Down Expand Up @@ -189,12 +189,20 @@ def kernel(x):
# Check the filename is reported correctly
self.assertIn('in the compilation of "error.cu"', msg)

@skip_with_cuda_python
# We need to run the test ina subprocess because the Linker class

@gmarkall gmarkall Feb 17, 2023

Copy link
Copy Markdown
Member

Choose a reason for hiding this comment

The reason will be displayed to describe this comment to others. Learn more.

Suggested change
# We need to run the test ina subprocess because the Linker class
# We need to run the test in a subprocess because the Linker class

# that instantiates either the CudaPythonLinker or CtypesLinker
# sets USE_NV_BINDING = config.CUDA_USE_NVIDIA_BINDING at
# module import time, so overriding the config variable does
# not help.

@TestCase.run_test_in_subprocess(envvars=_NUMBA_NVIDIA_BINDING_0_ENV)
def test_linking_cu_ctypes_unsupported(self):
link = os.path.join(os.path.dirname(__file__), 'data', 'jitlink.cu')
msg = ('Linking CUDA source files is not supported with the ctypes '
'binding')

with self.assertRaisesRegex(NotImplementedError, msg):
@cuda.jit('void()', link=['jitlink.cu'])
@cuda.jit('void()', link=[link])
def f():
pass

Expand Down