синхронизация массивов и многопоточность

Question

У меня есть многопоточное приложение, в котором я делаю следующее:

Thread1: читает данные с камеры и сохраняет их в массиве ArrayFire (бэкэнд CUDA).

Thread2: использование ArrayFire для вычисления определенных результатов на основе данных камеры.

Thread3: отправка результатов по сети. Вызывает функцию хоста в массиве.

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

При просмотре исходного кода ArrayFire я замечаю, что хост использует метод синхронизации потока CUDA, который (если я правильно понял) заставляет единственный поток CUDA, в котором все потоки выполняются fini sh все задачи, которые он выполняет.

Это заставляет Thread3 ждать, пока Thread2 завершит sh вычисление ArrayFire (если оно происходит в это время), и вызывает случайный джиттер в Thread3, где Мне нужно вызвать хост, чтобы скопировать массив в память ЦП.

Верны ли мои предположения и, если да, какие-либо предложения о том, что делать?

Answer 1

Выделите 4 указателя устройства и передайте два потоку 1 и два потоку 3. Это будет память, которую вы собираетесь использовать для отправки данных между ArrayFire и вашими потоками.

Создайте 2 потока, используя CUDA API. Один поток будет использоваться потоком 1 и один потоком 3. Также вызовите afcu :: getStream и получите поток, используемый ArrayFire.

Поток один собирается использовать свой поток для cudaMemcpyAsyn c для один из указателей устройства (убедитесь, что вы используете закрепленную память на хосте). Как только это будет сделано, вы запишете событие в свой поток и вызовете cudaStreamWaitEvent в потоке ArrayFire. Это скажет ArrayFire дождаться этого события, прежде чем двигаться дальше к вычислениям. На следующей итерации потока 1 запишите данные во второй указатель и т. Д.

В потоке ArrayFire. Вы будете вызывать af :: write для копии указателя на ваш входной массив. Вы можете рассматривать это как любой другой массив. Как только вы закончите, получите указатель устройства из массива результатов и вызовите cudaMemcpyAsyn c к указателю, указанному в потоке 3. Возможно, вы также захотите записать и дождаться событий здесь.

Это должно дать вам достаточное перекрытие между тремя нитями.

Answer 2

Я попробовал ваше предложение. Он не работает надежно и есть артефакты доступа к памяти.

Первый код:

входной поток:

cudaStream_t m_stream;
cudaEvent_t m_streamEvent;

cudaStreamCreateWithFlags(&m_stream, cudaStreamNonBlocking);
cudaEventCreate(&m_streamEvent);

int bytesPerPixel = 2;
int bytes = width * height * bytesPerPixel;

while(!stop)
{
    // Read from file to m_imageData

    // Next buufer is pinned memory allocated with af::pinned
    nextBuffer = getWriteBuffer();

    cudaMemcpyAsync(nextBuffer, m_imageData.data() + m_imageOffset, bytes, cudaMemcpyHostToDevice, m_stream);
    cudaEventRecord(m_streamEvent, m_stream);
    cudaStreamWaitEvent(m_stream, m_streamEvent, 0);   

    m_imageOffset =  (m_imageOffset + bytes) % m_imageData.size();
}

Процессорный поток:

tile x frames (batch) from ring buffer and push the result to a queue

Выходной поток:

cudaStream_t m_stream;
cudaEvent_t m_streamEvent;

cudaStreamCreateWithFlags(&m_stream, cudaStreamNonBlocking);
cudaEventCreate(&m_streamEvent);

int rgbaBufferSize = width * 4;
auto m_rgbaPinnedBuffer = af::pinned<quint8>(rgbaBufferSize);

while(!stop)
{
    rgba = m_queue.dequeue();
    rgbaAfBuffer = rgba.device<quint8>();
    cudaMemcpyAsync(m_rgbaPinnedBuffer, rgbaAfBuffer, rgbaBufferSize, cudaMemcpyDeviceToHost, m_stream);
    cudaEventRecord(m_streamEvent, m_stream);
    cudaStreamWaitEvent(m_stream, m_streamEvent, 0);
    rgba.unlock();

    // Do something with m_rgbaPinnedBuffer
}

При запуске с профилировщиком NVidia я вижу поток AF, я вижу поток выходного потока с выходами memcpy. Я не вижу поток входного потока, в котором есть memcpy. Я не знаю почему, хотя создание потока и события было сообщено успешно.

При использовании af :: host проблем с доступом к памяти нет, и я можно увидеть, что cudaMemcpyAsyn c происходит в потоке по умолчанию. Выходной RGBA выглядит следующим образом

При использовании cudaMemcpyAsyn c я вижу cudaMemcpyAsyn c на временной шкале потока, но иногда память повторяется. Это происходит чаще, когда я увеличиваю размер пакета или перемещаю другое приложение windows время быстрой кражи GPU. См. Вывод RGBA

Вы сталкивались с такой проблемой?