Javier Luis Lopez Ответов: 1

Барьеры Opencl не работают


Я попробовал ввести отправленный код.
Моя идея состояла в том, чтобы получить частичные суммы входных данных по массиву rms, затем сделать барьеры (глобальные и локальные), чтобы дождаться, пока все rms[k] будут заполнены, а затем суммировать их все, чтобы получить значение носителя.
Я поместил несколько printf, чтобы посоветовать, есть ли ошибки в исчислении.
Я получил ошибки в printf warning-2, но не warning-1 и 3 из-за добавления всех данных, потому что некоторые ядра все еще не закончили вычислять частичные суммы.

Я не использовал локальную память до тех пор, пока максимальный локальный размер составляет 256, что намного меньше высоты=10000

Как я заставляю GPU ждать, пока я не вычислю все частичные суммы?

Что я уже пробовал:

У меня есть следующий код:
__kernel void hallaRMS2(
	__global float*  data,  //size=WIDTH*HEIGHT
	int WIDTH,
	int HEIGHT,
	__global double *rms //size=HEIGHT
)
{
        int k = get_global_id(0); //0..HEIGHT
		__global float *x=data+k*WIDTH;


		double sum=0.0;

		for (int j=0;j<WIDTH;j++)
		{
			sum+=x[j];
		}

		rms[k]=sum;//to be used to calculate media
		if ((rms[k]<100*WIDTH)||(rms[k]>101*WIDTH)) printf("Warning-1: rms[%i]=%lg\n",k,rms[k]);

		barrier(CLK_GLOBAL_MEM_FENCE ); //to give time to all rms[k] be filled
		barrier(CLK_LOCAL_MEM_FENCE ); 
		

		if (k==0)
		{
			sum=0.0;
			for (int j=0;j<HEIGHT;j++)
			{
				if ((rms[j]<100*WIDTH)||(rms[j]>101*WIDTH)) printf("Warning-2: rms[%i]=%lg\n",j,rms[j]); 
				sum+=rms[j];
			}
			rms[0]=sum/(double) WIDTH/(double) HEIGHT;
			printf("GPU sum=%lg\n",sum);
			printf("GPU media=%lg\n",rms[0]);
		}
		else
			if ((rms[k]<100*WIDTH)||(rms[k]>101*WIDTH)) printf("Warning-3: rms[%i]=%lg\n",k,rms[k]);
...

1 Ответов

Рейтинг:
0

Javier Luis Lopez

Я не очень доволен этим решением:
Opencl 1.2 не позволяет синхронизировать все рабочие группы, как я уже говорил, поэтому он должен выйти из ядра и войти в новую, чтобы использовать данные из всех рабочих элементов.

Если кто-то знает, как это сделать в новом стандарте openCL 2.x, я был бы признателен.

К счастью для CUDA boys, он позволяет синхронизировать все устройства, не выходя из ядра. Это должно быть учтено, если кто-то попытается перевести код с Cuda на Opencl!