CUDA 动态并行性延迟全局内存

Question

为了测试动态并行性，我编写了一个简单的代码并在 GTX1080 上使用以下命令对其进行编译。

nvcc -arch=sm_35   -dc dynamic_test.cu -o dynamic_test.o
nvcc -arch=sm_35   dynamic_test.o  -lcudadevrt -o dynamic_test

但是，输出不如预期。似乎传递给子内核的指针被取消引用。

#include <stdlib.h>
#include <stdio.h>
#include <cublas_v2.h>
#include <cuda_runtime_api.h>

__global__ void child(int *a, int *b, int *c){

        int tid = threadIdx.x;
        c[tid] = a[tid] + b[tid];
}


__global__ void Parent(int *a, int *b, int *c){

        int tid = threadIdx.x;
        const int n = 10;

        a[tid] = tid;
        b[tid] = 2*tid;
        c[tid] = -10;

        __syncthreads();
        cudaDeviceSynchronize();
        if (tid == 1){
          child<<<1,n>>>(a,b,c);
          cudaDeviceSynchronize();
        }
}


int main(){

        int *d_a, *d_b, *d_c;
        const int n = 10;
        int a[n],b[n],c[n],i;

        cudaMalloc((void**)&d_a,n*sizeof(int));
        cudaMalloc((void**)&d_b,n*sizeof(int));
        cudaMalloc((void**)&d_c,n*sizeof(int));

        Parent << < 1, n >>> (d_a,d_b,d_c);
        cudaDeviceSynchronize();

        cudaMemcpy(a,d_a,n*sizeof(int),cudaMemcpyDeviceToHost);
        cudaMemcpy(b,d_b,n*sizeof(int),cudaMemcpyDeviceToHost);
        cudaMemcpy(c,d_c,n*sizeof(int),cudaMemcpyDeviceToHost);

        for(i=0; i<n; i++){
           printf("a[%d] = %d\n",i,a[i]);
        }
        for(i=0; i<n; i++){
           printf("b[%d] = %d\n",i,b[i]);
        }
        for(i=0; i<n; i++){
           printf("c[%d] = %d\n",i,c[i]);
        }

        cudaFree(d_a); cudaFree(d_b); cudaFree(d_c);

        return 0;
}

这是输出：

a[0] = 1
a[1] = 0
a[2] = 4208446
a[3] = 0
a[4] = 0
a[5] = 0
a[6] = 0
a[7] = 0
a[8] = 0
a[9] = 0
b[0] = 3
b[1] = 0
b[2] = 4204323
b[3] = 0
b[4] = 4205312
b[5] = 0
b[6] = 4732449
b[7] = 0
b[8] = 4205680
b[9] = 0
c[0] = 194906208
c[1] = 32767
c[2] = 4204143
c[3] = 0
c[4] = 4205616
c[5] = 0
c[6] = 4732608
c[7] = 0
c[8] = 4231155
c[9] = 0

阅读编程指南，我应该能够将全局变量传递给子内核，而不会导致任何延迟。我不确定为什么输出不正确。我的最终目标是在内核中使用 cublas 库。朝这个方向提出的任何建议也会有所帮助。

Answer 1

从 cuda 7.5 切换到 cuda 8.0 解决了这个问题。