Как распараллелить длинный тройной цикл

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

  1. for (int k1=1;k1<=16;k1++)
  2.  for (int t1=0;t1<=NPackets;t1++)  //t=0...T
  3.   for (int t2=t1;t2<=NPackets;t2++)  //k=t...T
  4.   {
  5.     vzaimkorr[k1][t2-t1]=vzaimkorr[k1][t2-t1]+kans[vk[k1][1]][t1]*kans[vk[k1][2]][t2]; 
  6.   }

NPackets - число точек(их 21000000, можно конечно взять пачками по 75000 точек, если это упростит задачу)
vzaimkorr - двухмерный массив из 16 корреляционных функций по NPackets точек типа float. (16xNPackets)
vk - двухмерный массив индексов 16х2 (нужен для взаимнокорреляционных функций, это "комбинации" каналов kans)
kans - массивы данных для обработки (16хNPackets)

допустим k1 можно раскидать на 16 блоков
а с остальными циклами как быть? особенно с последним, который от t1 до NPackets. я слышал, что видеокарты сейчас вроде как циклы сами распараллеливают некоторые, будет ли это эффективно здесь?

Forums: 

фактически это вычисление

фактически это вычисление взаимнокорреляционных функций между каналами 1...4
а vk=[ [1,1],[1,2],[1,3],...[4,1],[4,2],[4,3],[4,4] ]

Я не понимаю проблемы совсем.

Я не понимаю проблемы совсем.

Правая же часть (которая произведение kans[.][] * kans[][]) - не зависит от порядка вычислений и вообще ни от чего, параллелить можно как хочется, просто на все вкусы.

Оптимально, понятное дело, так, чтобы два блока не лезли внутрь одного участка vzaimkorr, тогда можно без излишней атомарности (ее контроля) обойтись.

Т.е. разверните задачу по другому: vzaimkorr[i[j] = (сумма ряда) и все станет понятно.

Fatal Error

решил двигаться от простого к сложному, написал на примере hellogpu.py(на питоне завожу)
вот код

  1. import pycuda.driver as drv
  2. import pycuda.tools
  3. import pycuda.autoinit
  4. import numpy
  5. import numpy.linalg as la
  6. from pycuda.compiler import SourceModule
  7.  
  8. mod = SourceModule("""
  9. __global__ void multiply_them(float *dest, float *a, float *b)
  10. {
  11.  const int i = threadIdx.x;
  12.  for (int t1=1;t1<=81920;t1++)
  13.   {
  14.   dest[i] = a[t1]*a[t1]+b[1]; //fmaf(a[t1],a[t1],b[0]);
  15.   }
  16. }
  17. """)
  18.  
  19. multiply_them = mod.get_function("multiply_them")
  20.  
  21. a = numpy.random.randn(81921).astype(numpy.float32)
  22. b = numpy.random.randn(81921).astype(numpy.float32)
  23.  
  24. dest = numpy.zeros_like(a)
  25. multiply_them(
  26.         drv.Out(dest), drv.In(a), drv.In(b),
  27.         grid=(16,1),
  28.         block=(512,1,1))
  29.  
  30. print b[0]

Отрабатывает нормально, считает, но если размер массива ставить больше (например в 10 раз пробовал), то при выполнении кода появляется синий экран смерти :( из-за чего это может быть?

сделал вот так const int i =

сделал вот так

  1.  const int i = (threadIdx.x + blockIdx.x * blockDim.x)*1000;
  2.  for (int t1=0;t1<=999;t1++)
  3.   {
  4.   dest[i+t1] = a[i+t1]*a[i+t1]+dest[i+t1]; //fmaf(a[t1],a[t1],b[0]);
  5.   }

и массив на 8192000 элементов. все отрабатывает нормально. но почему-то если цикл делать длинный, то BSOD выдает все равно

кое что наваял, только ошибку

кое что наваял, только экран мигает, а потом ошибку пишет после того как добавил"const int i1 = threadIdx.y + blockIdx.y * blockDim.y;", добавил в ядро условие if и объявил сетку не 65535, а [65535,65535] , не могу понять что я не так делаю? :(

  1. import pycuda.driver as drv
  2. import pycuda.tools
  3. import pycuda.autoinit
  4. import numpy
  5. import numpy.linalg as la
  6. from pycuda.compiler import SourceModule
  7.  
  8. mod = SourceModule("""
  9. __global__ void multiply_them(float *dest, float *a, float *b)
  10. {
  11.  const int i = threadIdx.x + blockIdx.x * blockDim.x;
  12.  const int i1 = threadIdx.y + blockIdx.y * blockDim.y;
  13.         if (i+i1 < 33553921)
  14. dest[i1] = fmaf(a[i],a[i+i1],dest[i1]); //a[]*a[]+dest[]
  15. __syncthreads();
  16. }
  17. """)
  18.  
  19. multiply_them = mod.get_function("multiply_them")
  20.  
  21. a = numpy.random.randn(33553921).astype(numpy.float32)
  22. b = numpy.random.randn(33553921).astype(numpy.float32)
  23.  
  24. dest = numpy.zeros_like(a)
  25. print "run..."
  26. multiply_them(
  27.         drv.Out(dest), drv.In(a), drv.In(b),
  28.         grid=(65535,65535),
  29.         block=(512,1,1))

вот лог:

LaunchError: cuCtxPopCurrent failed: launch timeout
Error in sys.exitfunc:
PyCUDA WARNING: a clean-up operation failed (dead context maybe?)
cuMemFree failed: invalid context
PyCUDA WARNING: a clean-up operation failed (dead context maybe?)
cuMemFree failed: invalid context
PyCUDA WARNING: a clean-up operation failed (dead context maybe?)
cuMemFree failed: invalid context
Traceback (most recent call last):
File "C:\Python27\lib\atexit.py", line 24, in _run_exitfuncs
func(*targs, **kargs)
pycuda._driver.LaunchError: cuCtxPopCurrent failed: launch timeout
PyCUDA WARNING: a clean-up operation failed (dead context maybe?)
cuModuleUnload failed: invalid context
-------------------------------------------------------------------
PyCUDA ERROR: The context stack was not empty upon module cleanup.
-------------------------------------------------------------------
A context was still active when the context stack was being
cleaned up. At this point in our execution, CUDA may already
have been deinitialized, so there is no way we can finish
cleanly. The program will be aborted now.
Use Context.pop() to avoid this problem.
-------------------------------------------------------------------

This application has requested the Runtime to terminate it in an unusual way.
Please contact the application's support team for more information.