Асинхронные потоки в CUDA не дают увеличения производительности

Question

Я пытаюсь ускорить следующий бит кода CUDA, используя несколько потоков.

#define N (4096 * 4096) 
#define blockDimX 16 
#define blockDimY 16 

float domain1 [N]; 
float domain2 [N]; 

__global__ updateDomain1_kernel(const int dimX, const int dimY) { 
    // update mechanism here for domain1 
    // ... 
} 

__global__ updateDomain2_kernel(const int dimX, const int dimY) { 
    // update mechanism here for domain2, which is nearly the same 
    // ... 
} 

__global__ addDomainsTogether_kernel(float* domainOut, 
            const int dimX, 
            const int dimY) 
{ 
    // add domain1 and domain2 together and fill domainOut 
} 

void updateDomains(float* domainOut) { 
    dim3 blocks((dimX + blockDimX - 1)/blockDimX , (dimY + blockDimY- 1)/blockDimY); 
    dim3 threads(blockDimX, blockDimY); 

    updateDomain1_kernel<<<blocks, threads>>> (dimX, dimY); 
    updateDomain2_kernel<<<blocks, threads>>> (dimX, dimY); 
    addDomainsTogether_kernel<<<block, threads>>> (domainOut_gpu, dimX, dimY); 
    cudaMemcpy(domainOut, domainOut_gpu, N * sizeof(float), cudaMemcpyDeviceToHost); 
} 

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

void updateDomains(float* domainOut) { 
    dim3 blocks((dimX + blockDimX - 1)/blockDimX , (dimY + blockDimY- 1)/blockDimY); 
    dim3 threads(blockDimX, blockDimY); 

    cudaStream_t stream0, stream1; 
    cudaStreamCreate(&stream0); 
    cudaStreamCreate(&stream1); 

    updateDomain1_kernel<<<blocks, threads, 0, stream0>>> (dimX, dimY); 
    updateDomain2_kernel<<<blocks, threads, 0, stream1>>> (dimX, dimY); 
    cudaDeviceSynchronize(); 

    addDomainsTogether_kernel<<<block, threads>>> (domainOut_gpu, dimX, dimY); 
    cudaMemcpy(domainOut, domainOut_gpu, N * sizeof(float), cudaMemcpyDeviceToHost); 

    cudaStreamDestroy(stream0); 
    cudaStreamDestroy(stream1); 
} 

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

void updateDomains(float* domainOut) { 
    dim3 blocks((dimX + blockDimX - 1)/blockDimX , (dimY + blockDimY- 1)/blockDimY); 
    dim3 threads(blockDimX, blockDimY); 

    cudaStream_t stream0; 
    cudaStreamCreate(&stream0); 

    updateDomain1_kernel<<<blocks, threads, 0, stream0>>> (dimX, dimY); 
    updateDomain2_kernel<<<blocks, threads, 0, stream0>>> (dimX, dimY); 

    addDomainsTogether_kernel<<<block, threads0, stream0>>> (domainOut_gpu, dimX, dimY); 
    cudaMemcpy(domainOut, domainOut_gpu, N * sizeof(float), cudaMemcpyDeviceToHost); 

    cudaStreamDestroy(stream0); 
} 

Однако, опять-таки вряд ли какая-то разница в скорость работы. Во всяком случае, последний кажется самым быстрым. Что заставляет меня думать, что есть что-то о потоках CUDA, которые я не понимаю. Может кто-нибудь просветить меня о том, как ускорить этот код?

Answer 1

Увеличенный параллелизм только увеличивает вашу вычислительную пропускную способность, если вы не были, уже используя все доступные ядра. Если у вас уже есть достаточный параллелизм, это не поможет вам, кроме как увеличить накладные расходы на синхронизацию.