cuda动态共享内存与报错调试HandleError

东方耀

cuda动态共享内存与报错调试HandleError
如果你想在一个核函数中动态地申请多个数组时该怎么办呢？看附件的图片

1. #include <stdio.h>
   
2. #define N   10
   
3. #define HANDLE_ERROR( err ) (HandleError( err, __FILE__, __LINE__ ))
   
4. static void HandleError( cudaError_t err,
   
5.                          const char *file,
   
6.                          int line ) {
   
7.     if (err != cudaSuccess) {
   
8.         printf( "%s in %s at line %d\n", cudaGetErrorString( err ),
   
9.                 file, line );
   
10.         exit( EXIT_FAILURE );
    
11.     }
    
12. }
    
13. 
    
14. __global__ void add( int *a, int *b, int *c ) {
    
15.     int tid = threadIdx.x;    // this thread handles the data at its thread id
    
16.     //使用了未指定大小的extern数组语法来声明共享内存数组
    
17.     //注意中括号与extern说明符
    
18.     //当核函数被启动时，数组大小从第三个执行配置参数被隐式地确定
    
19.     extern __shared__ int sh[];  // cuda动态共享内存的申明
    
20.     int *x = (int *)sh;      //动态内存的大小在调用时候的 第三个参数上
    
21.    
    
22.     if (tid >= N)
    
23.         return;
    
24.     x[tid] = a[tid] + b[tid];
    
25.     // 这里用共享内存是多此一举 为了展示用法而已
    
26.     c[tid] = x[tid];
    
27. }
    
28. 
    
29. //cuda动态共享内存与报错调试HandleError
    
30. //如果你想在一个核函数中动态地申请多个数组时该怎么办呢？看附件的图片
    
31. 
    
32. int main( void ) {
    
33.     int a[N], b[N], c[N];
    
34.     int *dev_a, *dev_b, *dev_c;
    
35. 
    
36.     // allocate the memory on the GPU
    
37.     HANDLE_ERROR( cudaMalloc( (void**)&dev_a, N * sizeof(int) ) );
    
38.     HANDLE_ERROR( cudaMalloc( (void**)&dev_b, N * sizeof(int) ) );
    
39.     HANDLE_ERROR( cudaMalloc( (void**)&dev_c, N * sizeof(int) ) );
    
40. 
    
41.     // fill the arrays 'a' and 'b' on the CPU
    
42.     for (int i=0; i<N; i++) {
    
43.         a[i] = -i;
    
44.         b[i] = i * i;
    
45.     }
    
46. 
    
47.     // copy the arrays 'a' and 'b' to the GPU
    
48.     HANDLE_ERROR( cudaMemcpy( dev_a, a, N * sizeof(int),
    
49.                               cudaMemcpyHostToDevice ) );
    
50.     HANDLE_ERROR( cudaMemcpy( dev_b, b, N * sizeof(int),
    
51.                               cudaMemcpyHostToDevice ) );
    
52.     //cuda共享内存的动态分配 这里有三个参数:block_num,thread_num,共享内存大小
    
53.     //每个线程块中共享内存的大小必须在核函数第三个执行配置参数中指定(以字节为单位)
    
54.     add<<<1,N,N>>>( dev_a, dev_b, dev_c );
    
55. 
    
56.     // copy the array 'c' back from the GPU to the CPU
    
57.     HANDLE_ERROR( cudaMemcpy( c, dev_c, N * sizeof(int),
    
58.                               cudaMemcpyDeviceToHost ) );
    
59. 
    
60.     // display the results
    
61.     for (int i=0; i<N; i++) {
    
62.         printf( "%d + %d = %d\n", a[i], b[i], c[i] );
    
63.     }
    
64. 
    
65.     // free the memory allocated on the GPU
    
66.     HANDLE_ERROR( cudaFree( dev_a ) );
    
67.     HANDLE_ERROR( cudaFree( dev_b ) );
    
68.     HANDLE_ERROR( cudaFree( dev_c ) );
    
69. 
    
70.     return 0;
    
71. }
    

复制代码

zouqiqi

让天下人人学会人工智能！人工智能的前景一片大好！