from numba.cuda.cudadrv import nvvm from numba.cuda.testing import ( unittest, skip_on_cudasim, SerialMixin, skip_unless_conda_cudatoolkit, ) from numba.cuda.cuda_paths import ( _get_libdevice_path_decision, _get_nvvm_path_decision, _get_cudalib_dir_path_decision, get_system_ctk, ) has_cuda = nvvm.is_available() has_mp_get_context = hasattr(mp, 'get_context') class LibraryLookupBase(SerialMixin, unittest.TestCase): def setUp(self): ctx = mp.get_context('spawn') qrecv = ctx.Queue() qsend = ctx.Queue() self.qsend = qsend self.qrecv = qrecv self.child_process = ctx.Process( target=check_lib_lookup, args=(qrecv, qsend), daemon=True,
from numba.cuda.cudadrv import nvvm from numba.cuda.testing import skip_on_cudasim, SerialMixin from numba.core import types, utils import unittest original = "call void @llvm.memset.p0i8.i64(" \ "i8* align 4 %arg.x.41, i8 0, i64 %0, i1 false)" missing_align = "call void @llvm.memset.p0i8.i64(" \ "i8* %arg.x.41, i8 0, i64 %0, i1 false)" @skip_on_cudasim('libNVVM not supported in simulator') @unittest.skipIf(utils.MACHINE_BITS == 32, "CUDA not support for 32-bit") @unittest.skipIf(not nvvm.is_available(), "No libNVVM") class TestNvvmWithoutCuda(SerialMixin, unittest.TestCase): def test_nvvm_llvm_to_ptx(self): """ A simple test to exercise nvvm.llvm_to_ptx() to trigger issues with mismatch NVVM API. """ def foo(x): x[0] = 123 cukern = compile_kernel(foo, args=(types.int32[::1], ), link=()) llvmir = cukern._func.ptx.llvmir ptx = nvvm.llvm_to_ptx(llvmir) self.assertIn("foo", ptx.decode('ascii')) def test_nvvm_memset_fixup(self):
from __future__ import absolute_import, print_function, division from numba.cuda.compiler import compile_kernel from numba.cuda.cudadrv import nvvm from numba.cuda.testing import skip_on_cudasim from numba import unittest_support as unittest from numba import types, utils @skip_on_cudasim('libNVVM not supported in simulator') @unittest.skipIf(utils.MACHINE_BITS == 32, "CUDA not support for 32-bit") @unittest.skipIf(not nvvm.is_available(), "No libNVVM") class TestNvvmWithoutCuda(unittest.TestCase): def test_nvvm_llvm_to_ptx(self): """ A simple test to exercise nvvm.llvm_to_ptx() to trigger issues with mismatch NVVM API. """ def foo(x): x[0] = 123 cukern = compile_kernel(foo, args=(types.int32[::1],), link=()) llvmir = cukern._func.ptx.llvmir ptx = nvvm.llvm_to_ptx(llvmir) self.assertIn("foo", ptx.decode('ascii')) if __name__ == '__main__': unittest.main()
from numba.config import IS_WIN32, IS_OSX from numba.cuda.cudadrv import nvvm from numba.cuda.testing import ( unittest, skip_on_cudasim, SerialMixin, skip_unless_conda_cudatoolkit, ) from numba.cuda.cuda_paths import ( _get_libdevice_path_decision, _get_nvvm_path_decision, _get_cudalib_dir_path_decision, get_system_ctk, ) has_cuda = nvvm.is_available() has_mp_get_context = hasattr(mp, 'get_context') class LibraryLookupBase(SerialMixin, unittest.TestCase): def setUp(self): ctx = mp.get_context('spawn') qrecv = ctx.Queue() qsend = ctx.Queue() self.qsend = qsend self.qrecv = qrecv self.child_process = ctx.Process( target=check_lib_lookup, args=(qrecv, qsend), daemon=True,