Выходная переменная ядра ядра CUDA не изменяется

Question

Я пытаюсь передать объект в ядро. Этот объект имеет в основном две переменные, один действует как входной, а другой - как вывод ядра. Но когда я запускаю ядро, выходная переменная не изменяется. Но когда я добавляю другую переменную в ядро ​​и присваиваю значение выходного значения этой переменной, она неожиданно срабатывает для обоих из них.

Я читал в другом потоке (While loop fails in CUDA kernel), что компилятор может оценивать ядро ​​как пустые для оптимизации целей, если он не производит никакого вывода.

Возможно, что этот объект ввода/вывода, который я передаю как единственный аргумент ядра, каким-то образом не распознается компилятором в качестве вывода? И если это правда. Есть ли элегантный способ (я хотел бы избежать добавления другого аргумента ядра), например, параметр компиляции, который может помешать этому?

Это класс для этого объекта.

class Replica 
{ 
    public : 
     signed char gA[1024]; 
     int MA; 
    __device__ __host__ Replica(){ 
    } 
}; 

И это ядро, которое в основном представляет собой сокращение суммы.

__global__ void sumKerA(Replica* Rd) 
{ 
    int t = threadIdx.x; 
    int b = blockIdx.x; 

    __shared__ signed short gAs[1024]; 
    gAs[t] = Rd[b].gA[t]; 

    for (unsigned int stride = 1024 >> 1; stride > 0; stride >>= 1){ 
     __syncthreads(); 
     if (t < stride){ 
      gAs[t] += gAs[t + stride]; 
     } 
    } 
    __syncthreads(); 

    if (t == 0){ 
     Rd[b].MA = gAs[0]; 
    } 
} 

И, наконец, мой код хозяина.

int main() 
{ 
    // replicas - array of objects 
    Replica R[128]; 
    for (int i = 0; i < 128; ++i){ 
     for (int j = 0; j < 1024; ++j){ 
      R[i].gA[j] = 2*(rand() % 2) - 1; 
     } 
     R[i].MA = 0; 
    } 

    Replica* Rd; 

    cudaSetDevice(0); 

    cudaMalloc((void **)&Rd,128*sizeof(Replica)); 
    cudaMemcpy(Rd,R,128*sizeof(Replica),cudaMemcpyHostToDevice); 

    dim3 DimBlock(1024,1,1); 
    dim3 DimGridA(128,1,1); 

    sumKerA <<< DimBlock, DimGridA >>> (Rd); 
    cudaThreadSynchronize(); 

    cudaMemcpy(&R,Rd,128*sizeof(Replica),cudaMemcpyDeviceToHost); 
    // cudaMemcpy(&M,Md,128*sizeof(int),cudaMemcpyDeviceToHost); 
    for (int i = 0; i < 128; ++i){ 
     cout << R[i].MA << " "; 
    } 

    cudaFree(Rd); 

    return 0; 
} 

Answer 1

Основываясь на вашем коде сокращения, кажется, что вы намерены запустить 1024 потока на блок.

В этом случае, это неверно:

dim3 DimBlock(1024,1,1); 
dim3 DimGridA(128,1,1); 

sumKerA <<< DimBlock, DimGridA >>> (Rd); 

Первый параметр конфигурации ядра является размеры сетки. Второй параметр - это размер блока threadblock. Если вы хотите 1024 потоков на блок, при запуске 128 блоков, ваш запуск ядра должен выглядеть следующим образом:

sumKerA <<< DimGridA, DimBlock >>> (Rd); 

Если добавить proper cuda error checking в свой код, я ожидаю, что вы увидели бы сбой запуска ядра, потому что с помощью блока переменная (blockIdx.x) для индексации в массив Rd из 128 элементов будет индексироваться за пределами конца массива в вашем исходном случае.

Если вы изменяете объекты Replica, на которые указывает Rd в вашем ядре, то есть внешнее видимое состояние, поэтому любой код, который модифицирует эти объекты, не может быть «оптимизирован» компилятором.

отметить также, что cudaThreadSynchronize() осуждается в пользу cudaDeviceSynchronize() (они имеют такое же поведение.)