сокращение openCL и передача 2d массива - PullRequest
Новая
       16

сокращение openCL и передача 2d массива

7 голосов
/ 06 января 2012

Вот цикл, который я хочу преобразовать в openCL. 

for(n=0; n < LargeNumber; ++n) {    
    for (n2=0; n2< SmallNumber; ++n2) {
        A[n]+=B[n2][n];
    }                                                         
    Re+=A[n];       
}

А вот то, что у меня есть до сих пор, хотя, я знаю, что это неправильно, и некоторые вещи пропущены. 

__kernel void openCL_Kernel( __global  int *A,
                         __global  int **B,  
                         __global  int *C, 
                         __global _int64 Re,
                                   int D) 
{

int i=get_global_id(0);
int ii=get_global_id(1);

A[i]+=B[ii][i];

//barrier(..); ?

Re+=A[i];

}

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

1 Ответ

11 голосов
/ 07 января 2012

Относительно вашей проблемы с передачей двойных указателей: эта проблема обычно решается путем копирования всей матрицы (или того, над чем вы работаете) в один непрерывный блок памяти и, если блоки имеют различную длину, передавая другой массив, который содержитсмещения для отдельных строк (чтобы ваш доступ выглядел примерно так: B[index[ii]+i]).

Теперь для сокращения до Re: поскольку вы не упомянули, над каким устройством вы работаете, ясобираюсь взять на себя свой GPU.В этом случае я бы не стал делать сокращение в том же ядре, поскольку оно будет чертовски медленным, как вы его опубликовали (вам придется сериализовать доступ к Re через тысячи потоков (и доступ к * 1006).* тоже). Вместо этого я написал бы нужное ядро, которое суммирует все B[*][i] в A[i] и помещает сокращение от A в Re в другом ядре и делает это в несколько шагов, то есть вы используете ядро ​​сокращениякоторый работает с элементом n и сводит их к чему-то вроде n / 16 (или к любому другому числу). Затем вы итеративно вызываете это ядро ​​до тех пор, пока не получите один элемент, который является вашим результатом (я делаю это описание намеренно расплывчатым, так как вы сказали, что хотите понять, подумайте сами).

Как примечание: вы понимаете, что исходный код точно не имеет хорошего шаблона доступа к памяти? Предполагая, что B относительно большой (и оченьбольше, чем A из-за второго измерения) итерация внутреннего цикла по внешнему индексуЕшьте много кашмир.Это еще хуже при портировании на gpu, который очень чувствителен к когерентному доступу к памяти

Так что переупорядочение так может значительно увеличить производительность: 

for (n2=0; n2< SmallNumber; ++n2)
  for(n=0; n < LargeNumber; ++n)    
    A[n]+=B[n2][n];
for(n=0; n < LargeNumber; ++n)                                                 
  Re+=A[n];

Это особенно верно, если у вас естькомпилятор, который хорош в автовекторизации, так как он может векторизовать эту конструкцию, но вряд ли он сможет сделать это для исходного кода (и если он не может доказать, что A и B[n2] могут 'Если говорить о той же памяти, она не может превратить исходный код в это).

...