OpenCL 에러

2013.01.21 14:20 from Program Language/OpenCL

void boo()

{

if(get_local_id(2)!=0) return;

foo();

barrier(CLK_LOCAL_MEM_FENCE);

}


작동 안한다.

알아서 작동할 줄 알았는데....


void boo()

{

if(get_local_id(2)!=0) 

foo();

barrier(CLK_LOCAL_MEM_FENCE);

}


이렇게 하자.






atomic 함수는 shared-reg에 써도 느리다.


왠만하면 피하자!


branch divergence를 유발한다!







저작자 표시 비영리 변경 금지
신고

'Program Language > OpenCL' 카테고리의 다른 글

GPGPU Computing의 기본 개념.  (0) 2013.04.19
OpenCL 에러  (0) 2013.01.21
global memory replay overhead  (0) 2012.12.21
[CUDA] occupancy  (0) 2012.11.07
NVIDIA clEnqueueReadBuffer non-blocking bug(?)  (0) 2012.11.02
Intel opencl platform analyzer  (0) 2012.10.15
TAG IT, OpenCL
Posted by Leo 리오 트랙백 0 : 댓글 0

http://software.intel.com/sites/landingpage/opencl/user-guide/index.htm

http://software.intel.com/sites/landingpage/opencl/user-guide/Using_the_Intel(R)_Graphics_Performance_Analyzers_(Intel(R)_GPA)_Platform_Analyzer.htm



AMD OpenCL컴파일러가 병맛같기 때문에 Intel용 툴을 써봤다.





1. GPA TOOL 다운


http://software.intel.com/en-us/vcsource/tools/intel-gpa







2. 프로파일 파일 생성


2-1. 환경 변수 설정


Environment Variable

  Default Value

CL_GPA_CONFIG_ENABLE_API_TRACING

  True







CL_CONFIG_USE_GPA  = True






2-2. Intel GAP Monitor 실행





2-3. 커맨드 설정



Command Line에 실행 Full Path와 argument씀.

Working Folder 설정.




2-4. 프로파일 설정.


Manage Profiles...



a. Enable trigger (단축키로 실행시킬 수 있지만, 귀찮)

b. Apllication Time, >= 0 ( 시작부터)

c. Frame and Trace Capture (Frame은 생성이 안된다;)

d. Retrigger Delay : 1secs (1초마다 생성, 어플이 끝날 때 한번 하고 싶었지만 모르겠다...)




2-5. Run


아래 내문서 폴더에 생성됬다고 뜬다.


1초마다 생성되어?

여러개가 생성되었다.




<span style="background-color: rgb(124, 67, 177); color: rgb(255, 255, 255); ">에러가 날 때</span >





3. Intel opencl Platform Analyzer


맨 마지막 파일을 실행시키면








4. 분석하기















PS. 근데 위에 환경변수들을 켜놓으면

성능이 엄청 나빠진다.

- 해결











저작자 표시 비영리 변경 금지
신고

'Program Language > OpenCL' 카테고리의 다른 글

[CUDA] occupancy  (0) 2012.11.07
NVIDIA clEnqueueReadBuffer non-blocking bug(?)  (0) 2012.11.02
Intel opencl platform analyzer  (0) 2012.10.15
DeviceQuery  (0) 2012.10.08
OpenCL 포팅  (0) 2012.10.05
Release  (0) 2012.09.24
Posted by Leo 리오 트랙백 0 : 댓글 0

OpenCL 포팅

2012.10.05 19:21 from Program Language/OpenCL

1. C로 구현.


int j = 0;

for( int i=0 ; i<256 ; i ++)

{

j =  j * 3;

a[i] = j ;

}

이 코드는 a[i]에 3^i을 넣는 코드이다.

물론 j =  j * 3; 대신 a[i] = 3^i;를 써도 되지만 j * 3이 3^i 보다 light(빠름)다면 이렇게 쓸 것이다.



2. OpenMP로 구현.

우선 OpenMP로 포팅해서 data dependency를 없애주자.


int j = 0;

#pragma omp parallel

for( int i=0 ; i<256 ; i ++)

{

j =  j * 3;

a[i] = j ;

}


3. OpenMP Data Dependency 제거

이렇게 하고 인텔 composer같은 툴을 돌리면 

j끼리 디펜던시가 있다고 알려준다.


바꾸자.

#pragma omp parallel

for( int i=0 ; i<256 ; i ++)

{

int j =  3 ^ i;

a[i] = j ;

}

이제 j는 for안의 local variable이다.

디펜던시제거 하였다.


이렇게 하면 4코어이면 빠르면 3.7배 정도 성능 향상이 있을 것이다.


4. OpenCL 구현 - Workitem 1개

이제 Workitem 1개의 OpenCL로 구현해보자.


(편의상 pseudo 코드)

for( int i=0 ; i<256 ; i ++)

{

write_buffer(a[], clmem); //인풋 넣기

executeKernel(1); //Workitem 1개로 실행

read_buffer(clmem, a[]); //결과값 빼오기.

}


__kernel execute(a[], i) //i값도 인풋으로 넣자.
{

int j =  3 ^ i;

a[i] = j ;

}

여기서 write_buffer와 read_buffer를 한번의 iteration마다 전체 배열a를 넣다 빼준다.
계산은 i번째 하나만 해준다.
그걸 for로 256번 돈다.
결과는 되게 느리다.

더보기



5. OpenCL 구현 - Workitem Max
앞의 구현된 코드의 결과가 맞다면 이제 Workitem 갯수를 늘려서 GPU의 많은 코어를 활용해보자.

//for( int i=0 ; i<256 ; i ++)

{

write_buffer(a[], clmem); //인풋 넣기

executeKernel(256); //Workitem 256개로 실행

read_buffer(clmem, a[]); //결과값 빼오기.

}


__kernel execute(a[]
{
int i =  get_global_id(0); // i 대신  get_global_id(0)를 사용하자.

int j =  3 ^ i;

a[i] = j ;

}


끝.











저작자 표시 비영리 변경 금지
신고

'Program Language > OpenCL' 카테고리의 다른 글

Intel opencl platform analyzer  (0) 2012.10.15
DeviceQuery  (0) 2012.10.08
OpenCL 포팅  (0) 2012.10.05
Release  (0) 2012.09.24
clCreateBuffer clEnqueueWriteBuffer clSetKernelArg clEnqueueReadBuffer  (0) 2012.09.24
OpenCL 외부파일 쉽게 컴파일하기  (0) 2012.09.21
TAG OpenCL
Posted by Leo 리오 트랙백 0 : 댓글 0



http://www.khronos.org/registry/cl/sdk/1.0/docs/man/xhtml/clCreateBuffer.html

cl_mem clCreateBuffer (cl_context contextcl_mem_flags flagssize_t sizevoid *host_ptrcl_int *errcode_ret)

flags
CL_MEM_READ_WRITE
CL_MEM_WRITE_ONLY
CL_MEM_READ_ONLY
CL_MEM_USE_HOST_PTR
CL_MEM_ALLOC_HOST_PTR
CL_MEM_COPY_HOST_PTR


cl_mem clEnqueueWriteBuffer (cl_command_queue command_queuecl_mem buffercl_bool blocking_writesize_t offsetsize_t cb, const void *ptrcl_uint num_events_in_wait_list, const cl_event *event_wait_listcl_event *event)


cl_int clSetKernelArg (cl_kernel kernelcl_uint arg_indexsize_t arg_size, const void *arg_value)


cl_int clEnqueueReadBuffer (cl_command_queue command_queuecl_mem buffercl_bool blocking_readsize_t offsetsize_t cbvoid *ptrcl_uint num_events_in_wait_list, const cl_event *event_wait_listcl_event *event)



NOTICE! cl_mem 자체가 포인터이다.
(typedef struct _cl_mem *            cl_mem;)


호스트 

cl_mem clmem_input = clCreateBuffer(context, CL_MEM_WRITE_ONLY, SIZE, 0, err);
clEnqueueWriteBuffer(commandq, clmem_input, CL_TRUE, 0, SIZE, INPUT_ptr, 0, 0);
clSetKernelArg(kernel, arg++, sizeof(cl_mem), clmem_input);

cl_mem clmem_result = clCreateBuffer(context, CL_MEM_READ_ONLY, SIZE, 0, err);
clSetKernelArg(kernel, arg++, sizeof(cl_mem),  clmem_result );

커널
__kernel Goo(global unsigned char* clmem1, local unsigned char* clmem2)
{

}

호스트
clEnqueueReadBuffer(commandq, clmem_result, CL_TRUE, 0, SIZE, OUTPUT_ptr, 0,0,0);

clReleaseMemObject(clmem_input);
clReleaseMemObject(clmem_result);




저작자 표시 비영리 변경 금지
신고
Posted by Leo 리오 트랙백 0 : 댓글 0

OpenCL에서 cl파일을 컴파일하기 위해선


파일을 string으로 읽어와야한다.


왜 이렇게 만들었지;


#include를 이용하면 간단하게 할 수 있다.


 

        char source2[512];

        char* ptr_src2 = source2;

        sprintf(source2, "#include \"%s\"\n",filename);

       

        program = clCreateProgramWithSource(context, 1, &ptr_src2, 0, &err);

 

        char build_option2[512] = "-I . ";

        err = clBuildProgram(program, 0, NULL, build_option2, NULL, NULL);

 


중요한점은 build_option에 "-I . "를 추가해줘야 include가 제대로 된다.


더보기


저작자 표시 비영리 변경 금지
신고
Posted by Leo 리오 트랙백 0 : 댓글 0

INTEL OPENCL 기준


#include 해서 같은 폴더에 넣어 놓더라도  

cannot open source file "......cl" 이라는 에러만 나고

include가 안된다.

Default로 어디를 참조하는지는 모르겠지만,

Build 할때 "-I ." 옵션을 주면 된다.

clBuildProgram(program, 0, NULL, build_option, NULL, NULL);


저작자 표시 비영리 변경 금지
신고
Posted by Leo 리오 트랙백 0 : 댓글 0