CUDA - 大数组初始化

Question

CUDA - 大数组初始化

ccuda

3

如何高效地初始化一个大的整数数组以供gpu使用？我需要给前两个元素赋值为1，其余为0（用于Eratosthenes筛法）。以下是几种可能的方法：

使用cudaMemcpy函数
使用cudaMemset函数加上在kernel中设置前两个元素的值
直接在kernel中进行初始化
其他方式

注意：数组大小是动态的（n是传递的参数）。

我的当前版本：

int array = (int*) malloc(array_size);
array[0] = 1;
array[1] = 1;
for (int i = 2; i < n; i++) {
    array[i] = 0;
}
HANDLE_ERROR(cudaMemcpy(dev_array, array, array_size, cudaMemcpyHostToDevice));
kernel<<<10, 10>>>(dev_array);

我需要一个例子，感激不尽。

- Bakus123

你有没有考虑使用calloc(它会将元素全部置零)而不是使用malloc，然后再将前两个元素设为1(就像你现在所做的一样)？ - P.P

1

@Blue Moon，谢谢您的想法，但是在CUDA程序中不可用calloc操作。 - Bakus123

2

最好使用 cudaMemset 而不是循环。 - haccks

@haccks，好的，但前两个元素怎么办？在单线程中初始化它吗？ - Bakus123

1

对于前两个元素：array[0] = array[1] = 0;。然后使用 cudaMemset(&array[2], 0, array_size * sizeof(int));。 - haccks

1个回答

网页内容由stack overflow 提供, 点击上面的

可以查看英文原文，
原文链接

- Grzegorz Szpetkowski · Accepted Answer

一种可能的方法是，如果 __device__ 数组具有常量大小，则可以通过在文件范围内（即任何函数之外）添加以下声明来直接在 GPU 上初始化该数组：

__device__ int dev_array[SIZE] = {1, 1};

剩余的元素将被初始化为零（您可以查看PTX汇编以确保这一点）。

然后，它可以像这样在内核中使用：

__global__ void kernel(void)
{
    int tid = ...;
    int elem = dev_array[tid];
    ...
}

如果变量大小不确定，您可以将cudaMalloc()与cudaMemset()结合使用：

int array_size = ...;
int *dev_array;

cudaMalloc((void **) &dev_array, array_size * sizeof(int));
cudaMemset(dev_array, 0, array_size * sizeof(int));

然后将前两个元素设置为1：

int helper_array[2] = {1, 1};
cudaMemcpy(dev_array, helper_array, 2 * sizeof(int), cudaMemcpyHostToDevice);

从计算能力2.0开始，您还可以通过 malloc() 设备函数直接在内核中分配整个数组：



__global__ void kernel(int array_size)
{
    int *dev_array;
    int tid = ...;

    if (tid == 0) {
        dev_array = (int *) malloc(array_size * sizeof(int));
        if (dev_array == NULL) {
            ...
        }
        memset(dev_array, 0, array_size * sizeof(int));
        dev_array[0] = dev_array[1] = 1;  
    }
    __syncthreads();

    ...
}

请注意，来自不同块的线程不知道屏障同步。
来自CUDA C编程指南中的内容：
CUDA内核中的malloc()函数从设备堆中分配至少size字节，并返回分配的内存的指针，如果不存在足够的内存来满足请求，则返回NULL。返回的指针保证对齐到16字节边界。
不幸的是，calloc()函数没有实现，因此您需要手动设置它。分配的内存在CUDA上下文的生命周期内存在，但您可以从此或后续内核中显式调用free()：
通过malloc()由给定的CUDA线程分配的内存在CUDA上下文的生命周期内保持分配状态，或者直到通过调用free()明确释放为止。它可以被任何其他CUDA线程使用，甚至可以从后续内核启动中使用。

说了这么多，我对补充的cudaMemcpy()并不是很在意，因为只需要复制两个元素，而且很可能只占总执行时间的不到0.01%（很容易进行性能分析）。选择让你的代码更清晰的任何方式都可以。否则就是过早优化。