Проблема с большими размерами массива в CUDA

Question

Я знакомлюсь с CUDA, написав калькулятор точечного продукта.Я хотел протестировать его с массивами большого размера, чтобы провести исследование синхронизации, чтобы проверить два разных способа сбора векторной суммы.Однако, когда размер массива превышает 1024, я получаю ошибки.Я не уверен, откуда проблема.Карта GTX460M с 1,5 ГБ оперативной памяти.Я использую карту для отображения (это ноутбук).Кроме того, я не уверен, откуда возникла проблема.

Вот строка компиляции nvcc:

nvcc D:\Research\CUDA\TestCode\test_dotProduct_1.cu --use_fast_math --gpu-architecture sm_13 --compiler-bindir="D:\Programming\VisualStudio\2010express\VC\bin" --machine 32 -o multi_dot.exe

Кажется, у меня также возникают проблемы с компиляцией в 64-битной версии, но это еще одна проблема

Вот вывод длямассив размера 1024:
РАСЧЕТ ХОЗЯЙСТВА: 357389824.000000
РАСЧЕТ DEV PARA: 357389824.000000
РАСЧЕТ DEV SERI: 357389824.000000

Вот вывод для массива размера 2048:
HOST C: 2861214720.000000
РАСЧЕТ DEV PARA: -1. # INF00
РАСЧЕТ DEV SERI: -1. # INF00

Вот мой код:

 /*Code for a CUDA test project doing a basic dot product with doubles
*
*
*
*/
 #include <stdio.h>
 #include <cuda.h>


 __global__ void GPU_parallelDotProduct(double *array_a, double *array_b, double     *array_c){
     array_c[threadIdx.x] = array_a[threadIdx.x] * array_b[threadIdx.x];
 }

 __global__ void GPU_parallelSumVector(double *vector, double *sum, int base){
    sum[threadIdx.x + blockIdx.x] = vector[blockIdx.x + threadIdx.x * base] +         vector[blockIdx.x + threadIdx.x * base + 1];
 }

__global__ void GPU_serialSumVector(double *vector, double *sum, int dim){
     for(int i = 0; i < dim; ++i){
         sum[0] += vector[i];
     }
}

__host__ void CPU_serialDot(double *first, double *second, double *dot, int dim){
     for(int i=0; i<dim; ++i){
         dot[0] += first[i] * second[i];
     }
 }

__host__ void CPU_serialSetupVector(double *vector, int dim, int incrSize, int start){
     for(int i=0; i<dim; ++i){
         vector[i] = start + i * incrSize;
     }
 }

 int main(){
     //define array size to be used
     //int i,j;
     const int VECTOR_LENGTH = 2048;
           int SUM_BASE      = 2;
           int SUM_ROUNDS    = VECTOR_LENGTH / SUM_BASE;
           int ELEMENT_SIZE  = sizeof(double);
           //   int currentSize   = VECTOR_LENGTH;
     //arrays for dot product
     //host
     double *array_a                  = (double*) malloc(VECTOR_LENGTH * ELEMENT_SIZE);
     double *array_b                  = (double*) malloc(VECTOR_LENGTH * ELEMENT_SIZE);
     double *dev_dot_product_parallel = (double*) malloc(VECTOR_LENGTH * ELEMENT_SIZE);
     double *dev_dot_product_serial   = (double*) malloc(VECTOR_LENGTH * ELEMENT_SIZE);
     double  host_dot_product         = 0.0;

     //fill with values
     CPU_serialSetupVector(array_a, VECTOR_LENGTH, 1, 0);
     CPU_serialSetupVector(array_b, VECTOR_LENGTH, 1, 0);
     CPU_serialDot(array_a, array_b, &host_dot_product, VECTOR_LENGTH);

     //device
     double *dev_array_a;
     double *dev_array_b;
     double *dev_array_c;
     double *dev_dot_serial;
     double *dev_dot_parallel;
     //allocate cuda memory
     cudaMalloc((void**)&dev_array_a,      ELEMENT_SIZE * VECTOR_LENGTH);
     cudaMalloc((void**)&dev_array_b,      ELEMENT_SIZE * VECTOR_LENGTH);
     cudaMalloc((void**)&dev_array_c,      ELEMENT_SIZE * VECTOR_LENGTH);
     cudaMalloc((void**)&dev_dot_parallel, ELEMENT_SIZE * VECTOR_LENGTH);
     cudaMalloc((void**)&dev_dot_serial,   ELEMENT_SIZE * VECTOR_LENGTH);


     //copy to from host to device
     cudaMemcpy(dev_array_a, array_a, ELEMENT_SIZE * VECTOR_LENGTH,     cudaMemcpyHostToDevice);
     cudaMemcpy(dev_array_b, array_b, ELEMENT_SIZE * VECTOR_LENGTH,     cudaMemcpyHostToDevice);
     cudaMemcpy(dev_dot_parallel, &dev_dot_product_parallel, ELEMENT_SIZE, cudaMemcpyHostToDevice);
     cudaMemcpy(dev_dot_serial, &dev_dot_product_serial, ELEMENT_SIZE, cudaMemcpyHostToDevice);

     //perform CUDA dot product
     GPU_parallelDotProduct<<<1, VECTOR_LENGTH>>>(dev_array_a, dev_array_b, dev_array_c);

     //condense a second vector in serial to compare speed up of tree condensing
     GPU_serialSumVector<<<1,1>>>(dev_array_c, dev_dot_serial, VECTOR_LENGTH);

     //condense vector (parallel)
     for(int i=SUM_ROUNDS; i>1; i/=SUM_BASE){
         GPU_parallelSumVector<<<1,i>>>(dev_array_c, dev_array_c, SUM_BASE);
     }
     GPU_parallelSumVector<<<1,1>>>(dev_array_c, dev_array_c, SUM_BASE);


     //get computed product back to the machine
     cudaMemcpy(dev_dot_product_parallel, dev_array_c, VECTOR_LENGTH * ELEMENT_SIZE,    cudaMemcpyDeviceToHost);
     cudaMemcpy(dev_dot_product_serial, dev_dot_serial, VECTOR_LENGTH * ELEMENT_SIZE, cudaMemcpyDeviceToHost);

     FILE *output = fopen("test_dotProduct_1.txt", "w");
     fprintf(output, "HOST CALCULATION:     %f \n", host_dot_product);
     fprintf(output, "DEV PARA CALCULATION: %f \n", dev_dot_product_parallel[0]);
     fprintf(output, "DEV SERI CALCULATION: %f \n", dev_dot_product_serial[0]);
     /*
     fprintf(output, "VALUES OF DEV_ARRAY_C VEC: \n");
     for(int i=0; i<VECTOR_LENGTH; ++i){
        fprintf(output, "value %i is: %f \n", i, dev_dot_product_parallel[i]);
     }
     */
     free(array_a);
     free(array_b);
     //free(host_dot_product);
     cudaFree(dev_array_a);
     cudaFree(dev_array_b);
     cudaFree(dev_array_c);
     cudaFree(dev_dot_parallel);
     cudaFree(dev_dot_serial);

    return(0);
}        

Answer 1

Максимальное количество потоков для блока для вашей карты - 1024, поэтому вы получаете сообщение об ошибке (для некоторых старых карт - 512).Вам нужно либо разделить свои блоки, чтобы использовать несколько измерений (опять же ограничено 1024 в направлении x, y, z на карте), либо использовать более одного блока в вашей сетке.