为啥同时使用 numba.cuda 和 CuPy 从 GPU 传输数据这么慢?
Posted
技术标签:
【中文标题】为啥同时使用 numba.cuda 和 CuPy 从 GPU 传输数据这么慢?【英文标题】:Why it is so slow to transfer data from GPU when use numba.cuda and CuPy at the same time?为什么同时使用 numba.cuda 和 CuPy 从 GPU 传输数据这么慢? 【发布时间】:2020-07-09 23:13:26 【问题描述】:我从 Cupy 的文档中阅读了有关如何同时使用 cupy 和 numba 以及使用 cuda 加速代码的示例。 https://docs-cupy.chainer.org/en/stable/reference/interoperability.html
我写了一个类似的代码来测试它:
import cupy
from numba import cuda
import numpy as np
import time
@cuda.jit('void(float32[:], float32[:], float32[:])')
def add(x, y, out):
start = cuda.grid(1)
stride = cuda.gridsize(1)
for i in range(start, x.shape[0], stride):
out[i] = x[i] + y[i]
a = cupy.arange(10000000)
b = a * 2
out = cupy.zeros_like(a)
print("add function time consuming:")
s = time.time()
add(a, b, out)
e = time.time()
print(e-s)
s = time.time()
print("out[2]:")
print(out[2])
e = time.time()
print("the time of transfering out[2] out of GPU:")
print(e-s)
s = time.time()
new_OUT = a + b
print("new out[2] which only use cupy:")
print(new_OUT[2])
e = time.time()
print("the total time of running cupy addition and transfering new out[2] out of GPU:")
print(e-s)
输出是:
add function time consuming:
0.0019025802612304688
out[2]:
6
the time of transfering out[2] out of GPU:
1.5608515739440918
new out[2] which only use cupy:
6
the total time of running cupy addition and transfering new out[2] out of GPU:
0.002993345260620117
第一种情况下out[2]的调用怎么会这么慢?
我正在编写一些需要处理一些cupy数组和矩阵的函数。这些函数工作正常,但是在运行这些函数之后,当我需要进行一些修改时,甚至调用out.shape
之类的东西都非常慢(我的矩阵和数组非常庞大)。
我不确定这里发生了什么,因为 cupy 也使用 cuda,所以当我调用 a + b
时,它应该在 GPU 上运行,但是当我调用 out[2]
来检查 out[ 2]。但是第一种情况消耗超高。
【问题讨论】:
【参考方案1】:要了解您的代码输出,至少需要注意两点:
在 CUDA 中,通常配置内核启动以指示网格配置(块数,每个块的线程数)。 numba CUDA 内核启动通常会在内核参数之前的方括号中显示网格配置:
kernel_name[grid_configuration](kernel_arguments)
在 numba CUDA 中,省略方括号和网格配置在语法上是允许的,这具有[1,1]
(一个块,由一个线程组成)的网格配置的隐含含义。您的内核使用或多或少的任意网格配置,因为它使用grid-stride loop。然而,这并不意味着网格配置无关紧要。它对性能很重要。 [1,1]
的网格配置将提供令人沮丧的性能,并且永远不应该在性能很重要的 CUDA 内核启动中使用。因此,我们可以通过更改您的内核调用来纠正此问题,例如:
add[1024,256](a, b, out)
这将启动一个由 1024 个块组成的网格,每个块有 256 个线程。
在 CUDA 中,内核启动是异步的。这意味着启动内核的主机代码将启动启动,但不会等待内核完成。这同样适用于 numba CUDA 内核启动。因此,内核启动本身的时间测量通常会令人困惑。出于计时目的,这可以通过强制 CPU 线程在计时区域内等待来调整,直到内核完成。在 numba CUDA 中,我们可以这样做:
cuda.synchronize()
【讨论】:
感谢您提供完整的解释!以上是关于为啥同时使用 numba.cuda 和 CuPy 从 GPU 传输数据这么慢?的主要内容,如果未能解决你的问题,请参考以下文章
可以在用户创建的 numba CUDA 设备函数中调用 numba.cuda.random 设备函数吗?
使用 Numba 进行矩阵乘法时出现 CUDA 内存不足错误
即使对于巨型矩阵,NUMBA CUDA 也比并行 CPU 慢
什么是 numba.cuda.local.array() 的有效替代方案,它们不像通过 to_device() 传递许多参数那么麻烦?