0

Я пишу на CUDA c++ 10.0. У меня есть массив на девайсе, который мне нужно регулярно заполнять новыми случайными числами. Для этого я написал kernel, с использованием параллелизации и прочего. Для генерации случайных чисел я решил использовать библиотеку thrust::random.

Допустим, мне надо заполнить массив из 100 чисел, и я это разбиваю на условные 10 блоков по 10 потоков. Получается вот такой кернель:

__global__ void gen(float* a, thrust::minstd_rand rng)
{
  int index = threadIdx.x + blockIdx.x * blockDim.x;
  thrust::uniform_real_distribution<float> dist(-10,10);
  a[index] = dist(gen);
 }

Сам вызов функции выглядит так:

int main()
{
  
  // инициализация devA на девайсе
  // ...
  
  thrust::minstd_rand rgen(time(0)); // нужно каждый раз разное, поэтому инициализируем временем системы
  
  
  gen<<<bloks, threads>>>(devA, rgen); // собственно вызов

  // ...
  
}

Сначала вопрос, где размещаются переменные, объявленные в кернеле, сразу на девайсе? thrust::uniform_real_distribution<float> dist и int index уже на девайсе, или как то там внутри бегают между CPU/GPU?

И, собственно, сам вопрос. Данная функция уже написана неправильно, так каждый вызов этой функции копирует себе состояние rgen, соответственно получаем много повторяющихся значений. Так же не уверен, что создание dist внутри кернеля целесообразно. Как вариант избежать копирований - rng передавать в виде ссылки на этот самый генератор, но тогда он должен находиться на девайсе, а еще, скорее всего, появится race condition, когда все потоки обратятся к этому генератору чуть ли не одновременно. Помогите разобраться, что с этим можно сделать? Или может быть я вообще не в ту сторону это делаю?

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

2
  • В вопросе явно не хватает конкретики. ¿"огромные" - это сколько в байтах и как часто? ¿Случайные - это любые псевдослучайные или криптографически приемлимые? 12 авг 2021 в 15:37
  • @user7860670, огромные - это у меня есть, допустим, 10гб на девайсе, есть вот этот массив и еще 3-4 int\bool\float таких же размеров, и чем больше впихну, тем лучше. В общем, 1+миллион. Как часто - в одной итерации всего 4 примерно равнозначных цикла, обновление рандомных значений - один из. Итераций от 50 до трех-четырех дней работы программы. Случайные - любые псевдослучайные, что бы при каждом запуске программы результат немного менялся, этого вполне достаточно, "взломом" этой последовательности заниматься никто никогда не будет.
    – AnonimYs
    12 авг 2021 в 17:30

1 ответ 1

0

Пришел к варианту, (вроде как) решающему проблему. Вариант - создать массив генераторов, каждый из которых будет работать для своего потока. Да, для каждого "объекта" будет существовать собственный набор генераторов. Код:

__global__ void initGen(thrust::uniform_real_distribution<float>* dist, thrust::minstd_rand rgn, time_t seed)     // seed - берется извне из time(0)
{
  int index = threadIdx.x + blockIdx.x * blockDim.x;
  rgn[index] = thrust::minstd_rand(seed + index * 10000);         //index добавляется для различий в сидах разных потоков\генераторов.
  gen[index] = thrust::uniform_real_distribution<float> dist(-10,10);
 }

__global__ void doGen(float* a, thrust::uniform_real_distribution<float>* dist, thrust::minstd_rand rgn)     
{
  int index = threadIdx.x + blockIdx.x * blockDim.x;
  a[index] = (*dist)(gen[index]);
}

int main()
{
  
  // инициализация devA на девайсе
  // ...
  
  thrust::minstd_rand *devRgn;
  cudaMalloc((void**)&devRgn, sizeof(thrust::minstd_rand) * size); // size -> of array
  // аналогично для uniform_real_distribution 
  
  
  int seedTime = time(0);
  initGen<<<blocks, threads>>>(devDist, devRgen, seed); 

  for(/*loop*/){
    doGen<<<blocks, threads>>>(devA, devDist, devRgen);
    // ... do smth with devA
  }

  // ...
  // free/cudaFree
  
}

Собственно, все генераторы записываются на девайс (даже не так, они создаются и живут только там) и занимают там место. Насколько рационально это по использованию памяти - не знаю, sizeof() пишет 4 и 8 для minstd_rand и uniform_real_distribution соответственно. Это аналогично трем дополнительным float/int на один "объект". Насколько приемлемо по реальной эффективности - посмотрим.

Ваш ответ

Нажимая на кнопку «Отправить ответ», вы соглашаетесь с нашими условиями использования и подтверждаете, что прочитали и поняли наши политику конфиденциальности и нормы поведения.

Всё ещё ищете ответ? Посмотрите другие вопросы с метками или задайте свой вопрос.