Мне нужен алгоритм для вычисления суммы параллельных префиксов массива без использования общей памяти. И если нет другой альтернативы использованию разделяемой памяти, как лучше всего решить проблему конфликтов?
Сумма параллельных префиксов CUDAfy
comment
Это описано в GPU Gems 3.
- person harold   schedule 10.01.2016
Ответы (1)
Эта ссылка содержит подробный анализ последовательного и параллельного алгоритмов для параллельной суммы префиксов:
Параллельная сумма префиксов (сканирование) с помощью CUDA
Он также содержит фрагмент кода C для реализации алгоритма параллельного префикса и подробное объяснение того, как избежать конфликтов с общей памятью.
Вы можете либо перенести коды в CUDAfy, либо просто определить области C и использовать их как неуправляемый код из своего приложения. Но в коде CUDA C есть несколько ошибок. Я пишу исправленную версию кода в Cudafy.NET
[Cudafy]
public static void prescan(GThread thread, int[] g_odata, int[] g_idata, int[] n)
{
int[] temp = thread.AllocateShared<int>("temp", threadsPerBlock);//threadsPerBlock is user defined
int thid = thread.threadIdx.x;
int offset = 1;
if (thid < n[0]/2)
{
temp[2 * thid] = g_idata[2 * thid]; // load input into shared memory
temp[2 * thid + 1] = g_idata[2 * thid + 1];
for (int d = n[0] >> 1; d > 0; d >>= 1) // build sum in place up the tree
{
thread.SyncThreads();
if (thid < d)
{
int ai = offset * (2 * thid + 1) - 1;
int bi = offset * (2 * thid + 2) - 1;
temp[bi] += temp[ai];
}
offset *= 2;
}
if (thid == 0)
{
temp[n[0] - 1] = 0;
} // clear the last element
for (int d = 1; d < n[0]; d *= 2) // traverse down tree & build scan
{
offset >>= 1;
thread.SyncThreads();
if (thid < d)
{
int ai = offset * (2 * thid + 1) - 1;
int bi = offset * (2 * thid + 2) - 1;
int t = temp[ai];
temp[ai] = temp[bi];
temp[bi] += t;
}
}
thread.SyncThreads();
g_odata[2 * thid] = temp[2 * thid]; // write results to device memory
g_odata[2 * thid + 1] = temp[2 * thid + 1];
}
}
Вы можете использовать приведенный выше модифицированный код вместо того, что указан в ссылке.
person
Alchemist
schedule
10.01.2016