Declare and initialize array of arrays in CUDA

Question

I am trying to declare and initialize an array of arrays in CUDA. I am using the following piece of code:

int** h_array = (int**)malloc(num_of_arrays * sizeof(int*));
int** d_array;
cudaMallocHost((void**)&d_array, num_of_arrays * sizeof(int*));

for(size_t i = 0 ; i < num_of_arrays ; i++){
    cudaMallocHost(&h_array[i], array_size * sizeof(int));
}
for(size_t i = 0 ; i < num_of_arrays ; i++){
    cudaMemcpy(d_array[i], h_array[i], array_size * sizeof(int), cudaMemcpyHostToDevice);
}

int** h_array2 = (int**)malloc(num_of_arrays * sizeof(int*));

Please note that h_array2 is initialized correctly(it is an array of arrays, for which each array is initialized correctly). Then I try to do the following:

for(size_t i = 0 ; i < num_of_arrays ; i++){
    cudaMemcpy(d_array[i], h_arra2[i], array_size * sizeof(int), cudaMemcpyHostToDevice);
}

To sum up, I try to declare and initialize an array of arrays in the device memory. I know that I can not access device memory from host memory.

The code above seems not to work.

Could you please tell me what is wrong and help me? Thank you in advance.

Answer 1

Something like this should work:

// create intermediate host array for storage of device row-pointers
int** h_array = (int**)malloc(num_of_arrays * sizeof(int*));
// create top-level device array pointer
int** d_array;
cudaMalloc((void**)&d_array, num_of_arrays * sizeof(int*));
// allocate each device row-pointer, then copy host data to it
for(size_t i = 0 ; i < num_of_arrays ; i++){
    cudaMalloc(&h_array[i], array_size * sizeof(int));
    cudaMemcpy(h_array[i], h_array2[i], array_size * sizeof(int), cudaMemcpyHostToDevice);
}
// fixup top level device array pointer to point to array of device row-pointers
cudaMemcpy(d_array, h_array, num_of_arrays * sizeof(int*), cudaMemcpyHostToDevice);

Here is a question/answer discussing various 2D and 3D array allocation methods.