Имеют ли дочерние ядра, запускаемые ядром, тот же размер деформации, что и ядра, запускаемые хостом?

Question

Когда блок ядра запускается с хоста, он имеет размер деформации 32. Это то же самое для дочерних ядер, запускаемых с помощью динамического параллелизма? Мое предположение было бы да, но я не видел это в документах.

Главный вопрос, конечно, таков: стоит ли это того?

__global__ void kernel(const int * vec, float * outvec){
    int idx = threadIdx.x;
    float random_sum=0;
    for(int j=0; j<vec[idx]; j++){
        random_sum+=threadsafe_rand_uniform();
    }
    outvec[idx] = random_sum;
}

Хорошо, этот пример немного надуманный. Суть, однако, в том, что если у вас есть цикл разной длины от потока к потоку, заманчиво попробовать и динамически распараллелить его. Однако, если деформация по-прежнему равна 32, в конечном итоге вы будете тратить много процессоров на деформации неравномерных размеров. В этом конкретном примере вы можете сначала отсортировать данные, а затем отправить динамически распараллеливаемые индексы в одном ядре и индексы неправильной формы в другом.

Answer 1

Они действительно имеют одинаковый размер основы. Но это потому, что размер деформации фиксирован для видеокарты. Все ядра, работающие на одной видеокарте, будут иметь одинаковый размер деформации.

Сегодня практически все графические процессоры используют размер основы 32, но в будущем он может измениться.

Возможно, вы думали о количестве потоков в ядре, а не о размере деформации? Если это так, то нет, они не обязательно одинаковы. Вы запускаете новое ядро ​​с динамическим параллелизмом так же, как вы запускаете его с хоста:

<<<blocks, threads>>>threadsafe_rand_uniform();

Помните, что это не то же самое, что простой вызов функции устройства, чем вы сейчас и занимаетесь.

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