为什么这个numba.cuda查找表实现失败？

Posted 2021-04-21

我正在尝试实现一个转换，它在某个阶段有一个<1K的查找表。在我看来，它似乎不应该对现代显卡造成问题。

但是下面的代码失败并出现未知错误：

from numba import cuda, vectorize
import numpy as np

tmp = np.random.uniform( 0, 100, 1000000 ).astype(np.int16)
tmp_device = cuda.to_device( tmp )

lut = np.arange(100).astype(np.float32) * 2.5
lut_device = cuda.to_device(lut)

@cuda.jit(device=True)
def lookup(x):
    return lut[x]

@vectorize("float32(int16)", target="cuda")
def test_lookup(x):
    return lookup(x)

test_lookup(tmp_device).copy_to_host() # <-- fails with cuMemAlloc returning UNKNOWN_CUDA_ERROR

我在做什么反对numba.cuda的精神？

即使用以下简化代码替换lookup也会导致相同的错误：

@cuda.jit(device=True)
def lookup(x):
    return x + lut[1]

一旦发生此错误，我基本上不再能够使用cuda上下文了。例如，通过cuda.to_device分配一个新数组会导致：

numba.cuda.cudadrv.driver.CudaAPIError: [719] Call to cuMemAlloc results in UNKNOWN_CUDA_ERROR

---

运行于：4.9.0-5-amd64＃1 SMP Debian 4.9.65-3 + deb9u2（2018-01-04）

驱动程序版本：390.25

numba：0.33.0

答案

通过以粗体修改部分来修复上述代码：

@cuda.jit(device=True)
def lookup(x):
    lut_device = cuda.const.array_like(lut)
    return lut_device[x]

我运行了多种代码变体，包括简单地从这个内核中触摸查找表，但不使用它的输出。这与@talonmies断言UNKNOWN_CUDA_ERROR通常伴随无效指令一起发生，我认为可能存在导致问题的共享内存约束。

上面的代码使整个过程工作。但是，我仍然不明白为什么要以深刻的方式。

如果有人知道并了解原因，请随时为此答案做出贡献。