CUDA: copia de la memoria de la GPU 1 es más lento en multi-GPU

Question

Mi empresa tiene una configuración de dos GTX 295, por lo que un total de 4 GPU en un servidor, y tenemos varios servidores. Nos GPU 1 específicamente fue lento, en comparación con la GPU 0, 2 y 3, así que escribí un pequeño test de velocidad para ayudar a encontrar la causa del problema.

//#include <stdio.h>
//#include <stdlib.h>
//#include <cuda_runtime.h>
#include <iostream>
#include <fstream>
#include <sstream>
#include <string>
#include <cutil.h>

__global__ void test_kernel(float *d_data) {
    int tid = blockDim.x*blockIdx.x + threadIdx.x;
    for (int i=0;i<10000;++i) {
        d_data[tid] = float(i*2.2);
        d_data[tid] += 3.3;
    }
}

int main(int argc, char* argv[])
{

    int deviceCount;                                                         
    cudaGetDeviceCount(&deviceCount);
    int device = 0; //SELECT GPU HERE
    cudaSetDevice(device);


    cudaEvent_t start, stop;
    unsigned int num_vals = 200000000;
    float *h_data = new float[num_vals];
    for (int i=0;i<num_vals;++i) {
        h_data[i] = float(i);
    }

    float *d_data = NULL;
    float malloc_timer;
    cudaEventCreate(&start);
    cudaEventCreate(&stop); cudaEventRecord( start, 0 );
    cudaMemcpy(d_data, h_data, sizeof(float)*num_vals,cudaMemcpyHostToDevice);
    cudaMalloc((void**)&d_data, sizeof(float)*num_vals);
    cudaEventRecord( stop, 0 ); cudaEventSynchronize( stop ); cudaEventElapsedTime( &malloc_timer, start, stop );
    cudaEventDestroy( start );
    cudaEventDestroy( stop );


    float mem_timer;
    cudaEventCreate(&start);
    cudaEventCreate(&stop); cudaEventRecord( start, 0 );
    cudaMemcpy(d_data, h_data, sizeof(float)*num_vals,cudaMemcpyHostToDevice);
    cudaEventRecord( stop, 0 ); cudaEventSynchronize( stop ); cudaEventElapsedTime( &mem_timer, start, stop );
    cudaEventDestroy( start );
    cudaEventDestroy( stop );

    float kernel_timer;
    cudaEventCreate(&start);
    cudaEventCreate(&stop); cudaEventRecord( start, 0 );
    test_kernel<<<1000,256>>>(d_data);
    cudaEventRecord( stop, 0 ); cudaEventSynchronize( stop ); cudaEventElapsedTime( &kernel_timer, start, stop );
    cudaEventDestroy( start );
    cudaEventDestroy( stop );

    printf("cudaMalloc took %f ms\n",malloc_timer);
    printf("Copy to the GPU took %f ms\n",mem_timer);
    printf("Test Kernel took %f ms\n",kernel_timer);

    cudaMemcpy(h_data,d_data, sizeof(float)*num_vals,cudaMemcpyDeviceToHost);

    delete[] h_data;
    return 0;
}

Los resultados se

GPU0 cudaMalloc tomó 0.908640 ms Copiar a la GPU tomó 296.058777 ms Prueba del núcleo tomó 326.721283 ms

GPU1 cudaMalloc tomó 0.913568 ms Copiar a la GPU tomó 663.182251 ms Prueba del núcleo tomó 326.710785 ms

GPU2 cudaMalloc tomó 0.925600 ms Copiar a la GPU tomó 296.915039 ms Prueba del núcleo tomó 327.127930 ms

GPU3 cudaMalloc tomó 0.920416 ms Copiar a la GPU tomó 296.968384 ms Prueba del núcleo tomó 327.038696 ms

Como se puede ver, el cudaMemcpy a la GPU es también el doble de la cantidad de tiempo para GPU1. Esto es consistente entre todos nuestros servidores, siempre es GPU1 que es lento. Cualquier idea por qué esto puede ser? Todos los servidores ejecutan Windows XP.

Answer 1

Esto era un problema de controladores. La actualización a la última versión del controlador arregló

Answer 2

Esto puede ser un problema con el bus PCI, intentar cambiar las tarjetas en diferentes ranuras para ver si persiste el problema. Si esto es un problema, copiar todos los datos en la GTX295 a través de la ranura más rápido y utilizar la parte superior SLI copiarlo a la otra GPU (bus PCI lento).

Answer 3

Si puede utilizarse GDDR de la tarjeta de vídeo más rápida a la carga, a continuación, se puede hacer un dispositivo tansfer dispositivo en ancho de banda mucho mucho más altas, que podrían ayudar a eliminar el problema también. Además, revise su ancho de banda con la prueba de ancho de banda de NVidia para obtener algunos resultados físicos y prueba.

Buena suerte!

Answer 4

¿Usted está funcionando en una configuración de doble procesador? Hay un error en la corriente Tylersburg chipsets de tal manera que el ancho de banda de la ruta 86 (0) a la GPU (1) es más lento que el camino directo desde x 86 (0) a la GPU (0). Intel debería lanzar una nueva versión para corregir este error. Intenta bloquear su proceso de prueba a una CPU específica utilizando taskset y ver los resultados que obtiene.

lo que respecta Marcos