0

我有一个 CUDA 应用程序，我正在尝试使用常量内存。但是当我在 main 函数所在的同一个文件中编写内核时，只有常量内存中的数据在内核中被识别。否则，如果我在其他文件中声明内核函数，则常量内存变为 0 并且操作正常运行。我提供了一个简单的虚拟代码，可以更轻松地解释问题。该程序有一个 48x48 矩阵，分为 16x16 块，我在其中存储随机数 1 到 50。在内核内部，我将存储在常量内存中的数字添加到块中的每一行。代码如下：

头文件：

#include <windows.h>
#include <dos.h>
#include <stdio.h>
#include <conio.h>
#include <math.h>

#include <cuda.h>
#include <cuda_runtime.h>
#include <cutil.h>
#include <curand.h>
#include <curand_kernel.h>

__constant__ int test_cons[16];

__global__ void test_kernel_1(int *,int *);

主程序：

int main(int argc,char *argv[])
{   int *mat,*dev_mat,*res,*dev_res;
    int i,j;
    int test[16 ]   = {1,2,3,4,5,6,7,8,9,10,11,12,13,14,15,16};
    cudaMemcpyToSymbol(test_cons,test,16*sizeof(int));
    mat = (int *)malloc(48*48*sizeof(int));
    res = (int *)malloc(48*48*sizeof(int));
    memset(res,0,48*48*sizeof(int));

    srand(time(NULL));
    for(i=0;i<48;i++)
    {   for(j=0;j<48;j++)
        {   mat[i*48+j] = rand()%(50-1)+1;
            printf("%d\t",mat[i*48+j] );
        }
        printf("\n");
    }

    cudaMalloc((void **)&dev_mat,48*48*sizeof(int));
    cudaMemcpy(dev_mat,mat,48*48*sizeof(int),cudaMemcpyHostToDevice);
    cudaMalloc((void **)&dev_res,48*48*sizeof(int));

    dim3 gridDim(48/16,48/16,1);
    dim3 blockDim(16,16,1);

    test_kernel_1<<< gridDim,blockDim>>>(dev_mat,dev_res);

    cudaMemcpy(res,dev_res,48*48*sizeof(int),cudaMemcpyDeviceToHost);

    printf("\n\n\n\n");
    for(i=0;i<48;i++)
    {   for(j=0;j<48;j++)
        {   printf("%d\t",res[i*48+j] );
        }
        printf("\n");
    }

    cudaFree(dev_mat);
    cudaFree(dev_res);
    free(mat);
    free(res);
    exit(0);
}

内核功能：

__global__ void test_kernel_1(int *dev_mat,int* dev_res)
{
    int row = blockIdx.y*blockDim.y+threadIdx.y;
    int col = blockIdx.x*blockDim.x +threadIdx.x;

    dev_res[row*48+col] = dev_mat[row*48+col] + test_cons[threadIdx.x];
}

现在，当我在主程序文件中声明内核函数以及主程序时，常量内存值是正确的，否则如果它在不同的文件中，test_cons[threadIdx.x]值将变为 0。

我遇到了这个链接，它讨论了同样的问题，但我没有正确理解。如果有人能告诉我为什么会发生这种情况以及我需要做些什么来避免这个问题，那将非常有帮助。任何形式的帮助将不胜感激。谢谢。

4

2 回答 2

2

我最近在这里回答了一个类似的问题

CUDA 可以处理引用设备代码（入口点）或其他文件中的符号的代码，但它需要使用设备链接进行单独编译（如我上面给出的链接中所述和链接）。（并且单独的编译/链接需要 CC 2.0 或更高版本）

因此，如果您修改链接步骤，您可以将__constant__变量放在给定文件中，并从不同的文件中引用它。

如果不是（如果你没有指定单独的编译和设备链接），那么引用变量的设备代码、引用__constant__变量的宿主代码__constant__以及变量本身的定义/声明，都需要在同一个文件.

所以这：

__constant__ int test_cons[16];

这：

cudaMemcpyToSymbol(test_cons,test,16*sizeof(int));

还有这个：

dev_res[row*48+col] = dev_mat[row*48+col] + test_cons[threadIdx.x];

都需要在同一个文件中。

于 2013-09-23T03:51:27.577 回答
1

上面的答案是完全可以接受的，我添加了这个，因为用户无法让它工作。您可以接受以上答案，仅供参考。

内核.cu 文件：

#include <stdio.h>

__constant__ int test_cons[16];

void copymemory (int *test)
{
      cudaMemcpyToSymbol(test_cons,test,16*sizeof(int));
}

__global__ void test_kernel_1(int *dev_mat,int* dev_res)
{
    int row = blockIdx.y*blockDim.y+threadIdx.y;
    int col = blockIdx.x*blockDim.x +threadIdx.x;

    if (threadIdx.x ==0)
    {
        printf ("testcons[0] is %d\n", test_cons[threadIdx.x]) ;
    }
    dev_res[row*48+col] = dev_mat[row*48+col] + test_cons[threadIdx.x];
 }

简单的.cu 文件

#include <stdio.h>
#include <math.h>

#include <cuda.h>
#include <cuda_runtime.h>
#include <curand.h>
#include <curand_kernel.h>


void copymemory (int *temp) ;

__global__ void test_kernel_1(int *,int *);

int main(int argc,char *argv[])
{
    int *mat,*dev_mat,*res,*dev_res;
    int i,j;
    int test[16 ]   = {1,2,3,4,5,6,7,8,9,10,11,12,13,14,15,16};
    mat = (int *)malloc(48*48*sizeof(int));
    res = (int *)malloc(48*48*sizeof(int));
    memset(res,0,48*48*sizeof(int));


    copymemory (test) ;
    srand(time(NULL));
    for(i=0;i<48;i++)
    {
            for(j=0;j<48;j++)
            {
                mat[i*48+j] = rand()%(50-1)+1;
                //printf("%d\t",mat[i*48+j] );
            }
            //printf("\n");
    }

    cudaMalloc((void **)&dev_mat,48*48*sizeof(int));
    cudaMemcpy(dev_mat,mat,48*48*sizeof(int),cudaMemcpyHostToDevice);
    cudaMalloc((void **)&dev_res,48*48*sizeof(int));
    dim3 gridDim(48/16,48/16,1);
    dim3 blockDim(16,16,1);

    test_kernel_1<<< gridDim,blockDim>>>(dev_mat,dev_res);
    cudaMemcpy(res,dev_res,48*48*sizeof(int),cudaMemcpyDeviceToHost);
    for(i=0;i<48;i++)
    {
           for(j=0;j<48;j++)
            {
                    //   printf("%d\t",res[i*48+j] );
            }
            //printf("\n");
    }

    cudaFree(dev_mat);
    cudaFree(dev_res);
    free(mat);
    free(res);
    exit(0);
}

我已经评论了你的 printf。内核中的 printf 打印值 1。我还通过更改 main 函数中的 test[0] 的值进行了测试，它运行良好。

于 2013-09-23T04:46:36.853 回答