Нужна ли копия DeviceToDevice в программе CUDA?

Я делаю следующие две операции:

  1. Добавление двух массивов => a + b = AddResult
  2. Умножение двух массивов => AddResult * a = MultiplyResult

В приведенной выше логике AddResult является промежуточным результатом и используется в качестве входных данных в следующей операции умножения.

#define N 4096         // size of array

__global__ void add(const int* a, const int* b, int* c)
{
    int tid = blockIdx.x * blockDim.x + threadIdx.x;
    if (tid < N) 
    {
        c[tid] = a[tid] + b[tid];
    }
}

__global__ void multiply(const int* a, const int* b, int* c)
{
    int tid = blockIdx.x * blockDim.x + threadIdx.x;
    if (tid < N) 
    {
        c[tid] = a[tid] * b[tid];
    }
}

int main() 
{
    int T = 1024, B = 4;            // threads per block and blocks per grid
    int a[N], b[N], c[N], d[N], e[N];
    int* dev_a, * dev_b, * dev_AddResult, * dev_Temp, * dev_MultiplyResult;

    cudaMalloc((void**)&dev_a, N * sizeof(int));
    cudaMalloc((void**)&dev_b, N * sizeof(int));
    cudaMalloc((void**)&dev_AddResult, N * sizeof(int));
    cudaMalloc((void**)&dev_Temp, N * sizeof(int));
    cudaMalloc((void**)&dev_MultiplyResult, N * sizeof(int));

    for (int i = 0; i < N; i++) 
    {    
        // load arrays with some numbers
        a[i] = i;
        b[i] = i * 1;
    }

    cudaMemcpy(dev_a, a, N * sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_b, b, N * sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_AddResult, c, N * sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_Temp, d, N * sizeof(int), cudaMemcpyHostToDevice);
    cudaMemcpy(dev_MultiplyResult, e, N * sizeof(int), cudaMemcpyHostToDevice);

    //ADD
    add << <B, T >> > (dev_a, dev_b, dev_AddResult);
    cudaDeviceSynchronize();

    //Multiply
    cudaMemcpy(dev_Temp, dev_AddResult, N * sizeof(int), cudaMemcpyDeviceToDevice); //<---------DO I REALLY NEED THIS?
    multiply << <B, T >> > (dev_a, dev_Temp, dev_MultiplyResult);
    //multiply << <B, T >> > (dev_a, dev_AddResult, dev_MultiplyResult);
    
    //Copy Final Results D to H
    cudaMemcpy(e, dev_MultiplyResult, N * sizeof(int), cudaMemcpyDeviceToHost);


    for (int i = 0; i < N; i++) 
    {
        printf("(%d+%d)*%d=%d\n", a[i], b[i], a[i], e[i]);
    }

    // clean up
    cudaFree(dev_a);
    cudaFree(dev_b);
    cudaFree(dev_AddResult);
    cudaFree(dev_Temp);
    cudaFree(dev_MultiplyResult);

    return 0;
}

В приведенном выше примере кода я передаю результаты сложения (т. е. dev_AddResult) в другой массив устройств (т. е. dev_Temp) для выполнения операции умножения.

ВОПРОС: Поскольку массив результатов сложения (т. е. dev_AddResult) уже находится на устройстве GPU, действительно ли мне нужно перенести его в другой массив? Я уже пытался запустить следующее ядро, напрямую предоставив dev_AddResult в качестве входных данных, и это дало те же результаты. Есть ли риск, связанный с прямой передачей вывода одного ядра в качестве ввода следующего ядра? Какие-нибудь передовые практики, которым нужно следовать?

да, вы можете использовать «выход» одного ядра в качестве «входа» для следующего. Почему бы просто не попробовать?

Robert Crovella 23.03.2022 14:28

@RobertCrovella: я уже пытался, и это сработало нормально, но я просто хотел знать, действительно ли это правильный путь и есть ли какие-либо риски?

skm 23.03.2022 14:35
Стоит ли изучать PHP в 2023-2024 годах?
Стоит ли изучать PHP в 2023-2024 годах?
Привет всем, сегодня я хочу высказать свои соображения по поводу вопроса, который я уже много раз получал в своем сообществе: "Стоит ли изучать PHP в...
Поведение ключевого слова "this" в стрелочной функции в сравнении с нормальной функцией
Поведение ключевого слова "this" в стрелочной функции в сравнении с нормальной функцией
В JavaScript одним из самых запутанных понятий является поведение ключевого слова "this" в стрелочной и обычной функциях.
Приемы CSS-макетирования - floats и Flexbox
Приемы CSS-макетирования - floats и Flexbox
Здравствуйте, друзья-студенты! Готовы совершенствовать свои навыки веб-дизайна? Сегодня в нашем путешествии мы рассмотрим приемы CSS-верстки - в...
Тестирование функциональных ngrx-эффектов в Angular 16 с помощью Jest
В системе управления состояниями ngrx, совместимой с Angular 16, появились функциональные эффекты. Это здорово и делает код определенно легче для...
Концепция локализации и ее применение в приложениях React ⚡️
Концепция локализации и ее применение в приложениях React ⚡️
Локализация - это процесс адаптации приложения к различным языкам и культурным требованиям. Это позволяет пользователям получить опыт, соответствующий...
Пользовательский скаляр GraphQL
Пользовательский скаляр GraphQL
Листовые узлы системы типов GraphQL называются скалярами. Достигнув скалярного типа, невозможно спуститься дальше по иерархии типов. Скалярный тип...
1
2
38
1
Перейти к ответу Данный вопрос помечен как решенный

Ответы 1

Ответ принят как подходящий

Да, для случая, который вы показали, вы можете использовать «выход» одного ядра в качестве «входа» для следующего без какого-либо копирования. Вы уже сделали это и подтвердили, что это работает, поэтому я не буду приводить примеры. Изменения в любом случае тривиальны - устраните промежуточную операцию cudaMemcpy и используйте тот же указатель dev_AddResult вместо указателя dev_Temp при вызове ядра с несколькими вызовами.

Что касается «рисков», то мне ничего не известно о приведенном вами примере. Переходя от этого примера к, возможно, более общему использованию, вы хотели бы убедиться, что вычисления добавления выходных данных завершены, прежде чем использовать их где-то еще.

Ваш пример уже делает это избыточно, используя как минимум 2 механизма:

  • вмешательство cudaDeviceSynchronize() - это заставляет завершить ранее выпущенную работу
  • семантика потока — одно правило семантики потока заключается в том, что работа, переданная в конкретный поток, будет выполняться в порядке выдачи. Элемент B, выданный в поток X, не начнется до тех пор, пока не завершится ранее выданный элемент A в поток X.

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

В более общем случае, если вы выпустили ядро ​​добавления и умножения в отдельные потоки, то CUDA не дает никаких гарантий порядка выполнения, даже если вы «выпустили» ядро ​​умножения после ядра добавления.

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

Вопрос относительно вашего последнего абзаца: в случае использования разных потоков, каковы другие способы, кроме cudaDeviceSynchronize(), чтобы обеспечить синхронизацию, чтобы убедиться, что предыдущее ядро ​​завершено перед запуском следующего ядра?

skm 23.03.2022 15:11
cudaStreamWaitEvent()docs.nvidia.com/cuda/cuda-runtime-api/… на одном потоке и cudaEventRecord()docs.nvidia.com/cuda/cuda-runtime-api/… на другом потоке приходят на ум для синхронизации двух потоков. Также cudaStreamSynchronize() docs.nvidia.com/cuda/cuda-runtime-api/… для синхронизации хоста и одного потока GPU.
Sebastian 23.03.2022 15:19

Или — более сложно — Cuda Graphs developer.nvidia.com/blog/cuda-graphs или для конкретных случаев использования Dynamic Parallelism developer.nvidia.com/blog/…

Sebastian 23.03.2022 15:23

@Sebastian: Спасибо за предложение о графиках CUDA. Я изучаю это и выглядит довольно многообещающе для моей цели. Мне просто интересно, почему в моих трех разных книгах по CUDA вообще ничего не упоминается о графиках CUDA.

skm 23.03.2022 15:56

Графики Cuda (задачи) — это расширенная функция, которая была представлена ​​около 3 лет назад в CUDA 10.

Sebastian 23.03.2022 16:01

Другие вопросы по теме