CUDA: выделение 2D массива на GPU

Question

Я уже прочитал следующий поток , но я не смог заставить свой код работать.
Я пытаюсь выделить 2D-массив на GPU, заполнить его значениями и скопировать еговернуться к процессору.Мой код выглядит следующим образом:

__global__ void Kernel(char **result,int N)
{
    //do something like result[0][0]='a';
}
int N=20;
int Count=5;
char **result_h=(char**)malloc(sizeof(char*)*Count);
char **result_d; 
cudaMalloc(&result_d, sizeof(char*)*Count);
for(int i=0;i<Count;i++)
{
    result_h[i] = (char*)malloc(sizeof(char)*N);    
    cudaMalloc(&result_d[i], sizeof(char)*N); //get exception here
}

//call kernel
//copy values from result_d to result_h
printf("%c",result_h[0][0])//should print a

Как мне этого добиться?

Answer 1

Вы не можете манипулировать указателями устройств в коде хоста, поэтому вызов cudaMalloc внутри цикла завершается неудачно.Вероятно, вам следует просто выделить один непрерывный блок памяти, а затем рассматривать его как сплющенный 2D-массив.

Answer 2

Для выполнения простейших 2D-операций на графическом процессоре я бы рекомендовал вам рассматривать его как одномерный массив. cudaMalloc - блок размером w * h * sizeof (char). Вы можете получить доступ к элементу (i, j) через индекс j * w + i.

Кроме того, вы можете использовать cudaMallocArray , чтобы получить 2D-массив. Это лучше понимает локальность, чем линейная карта 2D-памяти. Вы можете легко привязать это к текстуре, например.

Теперь, с точки зрения вашего примера, причина, по которой он не работает, заключается в том, что cudaMalloc манипулирует указателем хоста для указания на блок памяти устройства. Ваш пример выделил структуру указателя для results_d на устройстве. Если вы просто измените вызов cudaMalloc для results_d на обычный malloc, он должен работать так, как вы изначально предполагали.

При этом, возможно, один из двух вариантов, которые я обрисовал выше, мог бы работать лучше с точки зрения простоты обслуживания кода.

Answer 3

В следующем примере кода выделяется двухмерный массив значений с плавающей точкой шириной × высотой и показано, как циклически проходить элементы массива в коде устройства [1]

// host code

float* devPtr; 

int pitch;

cudaMallocPitch((void**)&devPtr, &pitch, width * sizeof(float), height); 

myKernel<<<100, 192>>>(devPtr, pitch); 

// device code 
__global__ void myKernel(float* devPtr, int pitch) 

{ 
 for (int r = 0; r < height; ++r) { 

  float* row = (float*)((char*)devPtr + r * pitch); 

  for (int c = 0; c < width; ++c) { 
            float element = row[c]; } 
                                           } 
 }

В следующем примере кода выделяется ширина× массив CUDA высоты одного 32-битного компонента с плавающей запятой [1]

 cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<float>(); 
 cudaArray* cuArray; 
 cudaMallocArray(&cuArray, &channelDesc, width, height);

В следующем примере кода копируется двумерный массив в массив CUDA, выделенный в предыдущих примерах кода [1]:

cudaMemcpy2DToArray(cuArray, 0, 0, devPtr, pitch, width * sizeof(float), height, 
cudaMemcpyDeviceToDevice);

Следующий пример кода копирует массив памяти somehost в память устройства [1]:

float data[256]; 
int size = sizeof(data); 
float* devPtr; 
cudaMalloc((void**)&devPtr, size); 
cudaMemcpy(devPtr, data, size, cudaMemcpyHostToDevice);

Вы можете понять эти примеры и применять их в своих целях.

[1] NVIDIA CUDA Compute Unified Device Architecture

Answer 4

При таком распределении вы выделяете адреса, которые действительны в памяти ЦП. Значение адресов передается как число без проблем, но однажды в памяти устройства адрес char * не будет иметь значения.

Создайте массив из N * максимальной длины текста и еще один массив длиной N, в котором будет указана длина каждого слова.

Это немного сложнее, но если вы обрабатываете определенный текст (например, пароли) Я бы предложил вам сгруппировать его по длине текста и создать специализированное ядро ​​для каждой длины

template<int text_width>
__global__ void Kernel(char *result,int N)
{
    //pseudocode
    for i in text_width:
        result[idx][i] = 'a'
}

и в коде вызова ядра вы указываете:

switch text_length
case 16:
   Kernel<16> <<<>>> ()