전역 메모리 플로트 배열을 통해 굴러 다니는 OpenCl 커널 루프

Question

기본 병렬 프로그래밍 개념을 이해하지 못한다고 생각합니다. 아래의 커널은 내가 가지고있는 문제를 재현하는 단순한/고안된 예제이다. "점"의 모든 값을 사용하여 값을 계산하고이를 "블록"의 모든 항목에 할당하려고 시도합니다. 이러한 배열의 크기에 대한 제한을 적용하려고합니다. "블록"배열을 크게 종료 할 수는 있지만 ("clEnqueueNDRangeKernel 직후에 clFinish를 호출 한 후)"points "가 100,000 개 이상의 부동 소수점으로 채워지면"invalid command queue "오류가 발생합니다. 왜 당신이 도와 줄 수 있습니까?

__kernel void openClTesting (__global float *blocks, __global float *points, int pointsCount) 
    { 
     int globalId = get_global_id(0); 
     int count = 0; 
     for (int i = 0; i < pointsCount; i++) 
     { 
      count++; 
     } 
     blocks[globalId] = count; 
    }; 

일부 장치 정보 :

CL_DEVICE_LOCAL_MEM_SIZE = 49,152 
CL_DEVICE_GLOBAL_MEM_SIZE = 2,147,483,648 
CL_DEVICE_MAX_MEM_ALLOC_SIZE = 536,870,912 

호스트 코드 : I 통지

#include "stdafx.h" 
#include "CL\opencl.h" 
#include <iostream> 
#include <fstream> 
#include <string> 
#include <stddef.h> 
#include <stdlib.h> 
#include <stdio.h> 

#define NUM_POINTS 100000 
#define NUM_BLOCKS 100000000 

struct openClData 
{ 
cl_device_id deviceId = NULL; 
cl_uint numDevices; 
cl_uint numPlatforms; 
cl_int ret; 
cl_platform_id *platforms = NULL; 
cl_context context; 
cl_command_queue commandQueue; 
cl_program program; 
cl_kernel kernel; 
char* kernelCode; 
cl_uint kernelCodeSize; 
size_t globalItemSize; 
size_t localItemSize = 1; 
}; 


char* getKernelCode(); 
void printErrorLog(openClData oclData); 
void printRet(openClData oclData, int line); 
int countFileChars(const char *fileName); 

int _tmain(int argc, _TCHAR* argv[]) 
{ 
openClData oclData; 
oclData.globalItemSize = NUM_POINTS; 
oclData.kernelCode = getKernelCode(); 
std::cout << oclData.kernelCode << std::endl; 
oclData.kernelCodeSize = strlen(oclData.kernelCode); 

int numPoints = NUM_POINTS; 
int numBlocks = NUM_BLOCKS; 
cl_long localMemSize = 0, globalMemSize = 0, maxAllocMemSize = 0; 
float *blocks = new float[numBlocks]{0}; 
float *points = new float[numPoints]{0}; 

//prepare platform, device, context and command queue 
oclData.ret = clGetPlatformIDs(0, NULL, &oclData.numPlatforms); 
printRet(oclData, __LINE__); 
oclData.platforms = (cl_platform_id *)malloc(oclData.numPlatforms * sizeof(cl_platform_id)); 
oclData.ret = clGetPlatformIDs(oclData.numPlatforms, oclData.platforms, NULL); 
printRet(oclData, __LINE__); 
oclData.ret = clGetDeviceIDs(oclData.platforms[0], CL_DEVICE_TYPE_GPU, 1, &oclData.deviceId, &oclData.numDevices); 
printRet(oclData, __LINE__); 
oclData.context = clCreateContext(NULL, 1, &oclData.deviceId, NULL, NULL, &oclData.ret); 
printRet(oclData, __LINE__); 
oclData.commandQueue = clCreateCommandQueue(oclData.context, oclData.deviceId, 0, &oclData.ret); 
printRet(oclData, __LINE__); 
//prepare cl_mem objects 
cl_mem memObjBlocks = clCreateBuffer(oclData.context, CL_MEM_READ_WRITE, sizeof(float) * numBlocks, NULL, &oclData.ret); 
printRet(oclData, __LINE__); 
cl_mem memObjPoints = clCreateBuffer(oclData.context, CL_MEM_READ_WRITE, sizeof(float) * numPoints, NULL, &oclData.ret); 
printRet(oclData, __LINE__); 
oclData.ret = clEnqueueWriteBuffer(oclData.commandQueue, memObjBlocks, CL_TRUE, 0, sizeof(float) * numBlocks, blocks, 0, NULL, NULL); 
printRet(oclData, __LINE__); 
oclData.ret = clEnqueueWriteBuffer(oclData.commandQueue, memObjPoints, CL_TRUE, 0, sizeof(float) * numPoints, points, 0, NULL, NULL); 
printRet(oclData, __LINE__); 
//prepare program 
oclData.program = clCreateProgramWithSource(oclData.context, 1, (const char**)&oclData.kernelCode, (const size_t *)&oclData.kernelCodeSize, &oclData.ret); 
printRet(oclData, __LINE__); 
oclData.ret = clBuildProgram(oclData.program, 1, &oclData.deviceId, NULL, NULL, NULL); 
printRet(oclData, __LINE__); 
if (oclData.ret == CL_BUILD_PROGRAM_FAILURE) printErrorLog(oclData); 
oclData.kernel = clCreateKernel(oclData.program, "openClTesting", &oclData.ret); 
printRet(oclData, __LINE__); 
//set arguments 
oclData.ret = clSetKernelArg(oclData.kernel, 0, sizeof(cl_mem), &memObjBlocks); 
printRet(oclData, __LINE__); 
oclData.ret = clSetKernelArg(oclData.kernel, 1, sizeof(cl_mem), &memObjPoints); 
printRet(oclData, __LINE__); 
oclData.ret = clSetKernelArg(oclData.kernel, 2, sizeof(int), &numPoints); 
printRet(oclData, __LINE__); 
//run 
oclData.ret = clEnqueueNDRangeKernel(oclData.commandQueue, oclData.kernel, 1, NULL, &oclData.globalItemSize, &oclData.localItemSize, 0, NULL, NULL); 
printRet(oclData, __LINE__); 
oclData.ret = clFinish(oclData.commandQueue); 
printRet(oclData, __LINE__); 
oclData.ret = clEnqueueReadBuffer(oclData.commandQueue, memObjBlocks, CL_TRUE, 0, sizeof(float) * numBlocks, blocks, 0, NULL, NULL); 
printRet(oclData, __LINE__); 
oclData.ret = clFinish(oclData.commandQueue); 
printRet(oclData, __LINE__); 
//print some device info 
oclData.ret = clGetDeviceInfo(oclData.deviceId, CL_DEVICE_LOCAL_MEM_SIZE, sizeof(cl_ulong), &localMemSize, 0); 
std::cout << "CL_DEVICE_LOCAL_MEM_SIZE = " << localMemSize << '\n'; 
oclData.ret = clGetDeviceInfo(oclData.deviceId, CL_DEVICE_GLOBAL_MEM_SIZE, sizeof(cl_long), &globalMemSize, 0); 
std::cout << "CL_DEVICE_GLOBAL_MEM_SIZE = " << globalMemSize << '\n'; 
oclData.ret = clGetDeviceInfo(oclData.deviceId, CL_DEVICE_MAX_MEM_ALLOC_SIZE, sizeof(cl_long), &maxAllocMemSize, 0); 
std::cout << "CL_DEVICE_MAX_MEM_ALLOC_SIZE = " << maxAllocMemSize << '\n'; 

//clean up 
oclData.ret = clFlush(oclData.commandQueue); 
printRet(oclData, __LINE__); 
oclData.ret = clFinish(oclData.commandQueue); 
printRet(oclData, __LINE__); 
oclData.ret = clReleaseKernel(oclData.kernel); 
printRet(oclData, __LINE__); 
oclData.ret = clReleaseProgram(oclData.program); 
printRet(oclData, __LINE__); 
oclData.ret = clReleaseMemObject(memObjBlocks); 
printRet(oclData, __LINE__); 
oclData.ret = clReleaseMemObject(memObjPoints); 
printRet(oclData, __LINE__); 
oclData.ret = clReleaseCommandQueue(oclData.commandQueue); 
printRet(oclData, __LINE__); 
oclData.ret = clReleaseContext(oclData.context); 
printRet(oclData, __LINE__); 
for (size_t i = 0; i < 10; i++) 
{ 
    std::cout << blocks[i] << std::endl; 
} 
delete blocks; 
delete points; 
return 0; 
} 

char* getKernelCode() 
{ 
char* kernelCode = 
    "__kernel void openClTesting (__global float *blocks, __global float *points, int pointsCount)" 
    "{" 
    " int globalId = get_global_id(0);" 
    " int count = 0;" 
    " for (int i = 0; i < pointsCount; i++)" 
    " {" 
    "  count++;" 
    " }" 
    "blocks[globalId] = count;" 
    "}"; 
return kernelCode; 
} 

void printErrorLog(openClData oclData) 
{ 
size_t log_size; 
clGetProgramBuildInfo(oclData.program, oclData.deviceId, CL_PROGRAM_BUILD_LOG, 0, NULL, &log_size); 
char *log = (char *)malloc(log_size); 
clGetProgramBuildInfo(oclData.program, oclData.deviceId, CL_PROGRAM_BUILD_LOG, log_size, log, NULL); 
std::cout << log; 
free(log); 
} 

void printRet(openClData oclData, int line) 
{ 
std::cout << line << ", " << oclData.ret << std::endl; 
} 

int countFileChars(const char *fileName) 
{ 
std::ifstream ifs(fileName); 
ifs.seekg(0, std::ios_base::end); 
size_t count = ifs.tellg(); 
ifs.seekg(0, std::ios_base::beg); 
return count; 
} 

Answer 1

몇 가지 :

• 당신은 NUM_POINTS 작업 항목을 실행하고 있지만 각각의 결과를에 씁니다.- NUM_BLOCKS 개의 항목이 있습니다. NUM_POINTS가 NUM_BLOCKS보다 클 경우 정의되지 않은 동작입니다. 또한 NUM_BLOCKS를 변경하면 아무 것도하지 않는 이유에 대해 설명합니다 (위의 제한 외). 메모리 할당을 제외하고 NUM_BLOCKS의 값은 아무런 효과가 없습니다. (발견 한 메모리 할당 제한은 구현에 대한 CL_DEVICE_MAX_MEM_ALLOC_SIZE 값과 대략 일치합니다.)
• 여기에서 커널 시간 초과 조건으로 실행 중일 수 있습니다. 하나의 작업 항목에서 100000 루프 반복은 상당히 많이 발생합니다. OpenCL 구현에 따라 커널을 실행하는 데 너무 오래 걸리는 경우 커널을 종료 할 수 있습니다. 사용 가능한 스레드 병렬 처리를보다 잘 사용하고 루프가 아닌 작업 항목간에 작업을보다 수평 적으로 분리하십시오. 일반적으로 단기간 가동되는 많은 작업 항목이 장기 실행중인 것보다 낫습니다. 당신이 당신의 컴퓨팅 장치가 병렬로 실행할 수있는 작업 그룹의 수에 병렬 줄일 하나의 작업 항목으로 구성하는 모든 OpenCL을 작업 그룹을 강제하기 때문에
일반 노트에

Answer 2

, localItemSize = 1;은 피해야한다 이는 실행할 수있는 작업 항목 수보다 훨씬 적습니다. 당신이 NUM_POINTS 작업 그룹을 생성하기 때문에

clEnqueueNDRangeKernel(queue, kernel, 1, NULL, &globalSize, NULL, 0, NULL, NULL); 

이것은 또한 오류의 원천이 될 수 있지만 : 당신은 단순히 합리적인 자신의 가치를 OpenCL을 구현 그림을 가지고 대신 로컬 항목 크기에 대한 NULL를 전달할 수 있습니다 장치의 대기열 크기는 메모리 제한 (CL_DEVICE_QUEUE_ON_DEVICE_MAX_SIZE)입니다.