Синхронизация устройства во время асинхронной memcpy в CUDA - PullRequest
Новая
       18

Синхронизация устройства во время асинхронной memcpy в CUDA

4 голосов
/ 11 августа 2011

Предположим, я хочу выполнить асинхронный хост memcpy для устройства в CUDA, а затем немедленно запустить ядро. Как я могу проверить в ядре, завершена ли асинхронная передача?

1 Ответ

8 голосов
/ 12 августа 2011

Последовательность вашей асинхронной копии и запуска ядра с использованием «потока» CUDA гарантирует, что ядро ​​выполнится после завершения асинхронной передачи. Следующий пример кода демонстрирует: 

#include <stdio.h>

__global__ void kernel(const int *ptr)
{
  printf("Hello, %d\n", *ptr);
}

int main()
{
  int *h_ptr = 0;

  // allocate pinned host memory with cudaMallocHost
  // pinned memory is required for asynchronous copy
  cudaMallocHost(&h_ptr, sizeof(int));

  // look for thirteen in the output
  *h_ptr = 13;

  // allocate device memory
  int *d_ptr = 0;
  cudaMalloc(&d_ptr, sizeof(int));

  // create a stream
  cudaStream_t stream;
  cudaStreamCreate(&stream);

  // sequence the asynchronous copy on our stream
  cudaMemcpyAsync(d_ptr, h_ptr, sizeof(int), cudaMemcpyHostToDevice, stream);

  // sequence the kernel on our stream after the copy
  // the kernel will execute after the copy has completed
  kernel<<<1,1,0,stream>>>(d_ptr);

  // clean up after ourselves
  cudaStreamDestroy(stream);
  cudaFree(d_ptr);
  cudaFreeHost(h_ptr);
}

И вывод: 

$ nvcc -arch=sm_20 async.cu -run
Hello, 13

Я не верю, что существует какой-либо поддерживаемый способ проверить изнутри ядра, было ли выполнено какое-либо асинхронное условие (такое как завершение асинхронной передачи). Предполагается, что блоки потоков CUDA выполняются полностью независимо от других потоков выполнения.

...