Programming/openMP2012. 6. 18. 23:00
2C4T 시스템이니 4의 배수로 한번 테스트

parallel for를 통해서 자동으로 나누어 주도록 하면 되는데,
결과를 보면 순서가 틀어져서 나온다. (나누어 실행하되 누가 먼저 실행될진 모르니)
$ cat test.c
#include <stdio.h>
#include <omp.h>

int main(int argc, const char *argv[])
{
    int idx = 0;
    #pragma omp parallel for
    for(idx = 0; idx < 16; idx++)
         printf("%d\n",idx);

    return 0;
} 

$ ./a.out
0
1
2
3
12
13
14
15
4
5
6
7
8
9
10
11 

'Programming > openMP' 카테고리의 다른 글

TBB - Threading Building Blocks by intel  (0) 2013.01.08
openMP 문서들  (0) 2012.06.18
libgomp 공식 사이트 / 문서  (0) 2012.06.10
우분투에서 openMP 예제  (0) 2012.06.09
openMP / gcc -fopenmp  (0) 2012.06.09
Posted by 구차니
Programming/C Win32 MFC2012. 6. 14. 22:59
엔디안은 "메모리에 저장되는" 바이트의 순서를 의미한다.
"메모리에" 라는것이 매우 중요한데

HDD와 같은 녀석은 bit stream 처럼 어떻게 보면 big endian 처럼 저장되며
cpu 레지스터에서도 일단~은 big endian 처럼 보인다.

유일(?)하게 영향을 받는게
메모리에서 내용을 받아와 다른 형(type)으로 변환하는 경우인데
어셈블리 언어로 이야기 하자면 mem to register 명령에 영향을 미친다고 하면 되려나?

아무튼 프로그래밍을 5년 넘게 현업으로 하고 있지만
정말 제대로 엔디안을 아는게 아닌게 아니었구나 라는 생각이 문든 드는 화두..
"비트 쉬프트 할 경우 정말 실제로는 어떻게 작동할 것인가?"
요 녀석에 멘붕을 느끼는중

[링크 : http://www.terms.co.kr/big-endian.htm ]
[링크 : http://ko.wikipedia.org/wiki/엔디언 ]

[링크 : http://www.ibm.com/developerworks/kr/library/au-endianc/index.html ] 
Posted by 구차니
Programming/C Win32 MFC2012. 6. 12. 14:14
부제 : 아오 미네랄 써글넘의 localtime()

localtime() 함수는 struct tm * 형을 리턴하는데
다르게 말하면, glibc나 library 내의 변수의 포인터를 리턴하는 식이 되는지라 매번 할당해서 돌려주는게 아니라는 의미.
즉, 연속으로 localtime을 사용해서 포인터로 받는다면, 당연히 동일 주소 동일 내용이 되므로
조건식 비교에서 항상 참이 될 수 밖에 없다 -_-

struct tm *localtime(const time_t *timep);
struct tm *localtime_r(const time_t *timep, struct tm *result);

POSIX.1-2001 says: "The asctime(), ctime(), gmtime(), and localtime() functions shall return values in one of two static objects: a broken-down time structure and an array of type char. Execution of any of the functions may overwrite the information returned in either of these objects by any of the other functions." This can occur in the glibc implementation.

[링크 : http://linux.die.net/man/3/localtime ]   


그런 이유로 아래와 같이 복사하거나, 처음부터 포인터가 아닌 값으로 받아 변수에 넣도록 해주는 것도 방법인데
오홍.. 아래 방법은 당연한 문법이지만 왜이리 생소해 보일까? ㅋㅋ

struct tm stTempTime;
pstCurTime = localtime(&lCurTime);
memcpy(&stTempTime, pstCurTime, sizeof(struct tm));

는 간단하게
struct tm stTempTime = *localtime(&lCurTime);

[링크 : http://kldp.org/node/71959 ]  

'Programming > C Win32 MFC' 카테고리의 다른 글

c 변수범위 헤더  (0) 2012.07.02
엔디안 / endian  (2) 2012.06.14
C언에 이스케이프 문자를 이용한 특수문자 출력하기  (0) 2012.03.28
함수 포인터 배열  (0) 2012.03.07
헐 # include 이게 되는거였다니!  (0) 2012.02.15
Posted by 구차니
Programming/openMP2012. 6. 10. 20:08
openMP의 GNU 구현 버전인 libgomp(Gnu OpenMP)

[링크 : http://gcc.gnu.org/onlinedocs/libgomp/]
[링크 : http://gcc.gnu.org/onlinedocs/libgomp.pdf]

'Programming > openMP' 카테고리의 다른 글

openMP 문서들  (0) 2012.06.18
openmp for문 나누기  (0) 2012.06.18
우분투에서 openMP 예제  (0) 2012.06.09
openMP / gcc -fopenmp  (0) 2012.06.09
openMP  (0) 2011.02.05
Posted by 구차니
Programming/openMP2012. 6. 9. 19:26
예제를 따라하고 출력을 해보니 먼가 이상한거 발견
hello가 아니라 hell world래.. 지옥에 오신걸 환영합니다 인가 -_-
atom330(2core / 4thread) 이라서 일단 4개 쓰레드로 기본 실행 된 듯.

$ vi test.c
#include <stdio.h>
#include <omp.h>

int main(int argc, const char *argv[])
{
    #pragma omp parallel
    printf("hell world\n");

    return 0;
}
 
$ gcc -fopenmp test.c
$ ./a.out
hell world
hell world
hell world
hell world

[링크 : http://assaart.co.cc/wordpress/?p=59

'Programming > openMP' 카테고리의 다른 글

openMP 문서들  (0) 2012.06.18
openmp for문 나누기  (0) 2012.06.18
libgomp 공식 사이트 / 문서  (0) 2012.06.10
openMP / gcc -fopenmp  (0) 2012.06.09
openMP  (0) 2011.02.05
Posted by 구차니
Programming/openMP2012. 6. 9. 19:14
openMP는 컴파일러의 도움을 받아야 하기 때문에
소스에 openMP를 사용하는지에 대한 플래그를 컴파일러에 넘겨주어야 한다.
gcc의 경우 -fopenmp를 통해서 openmp의 사용을 알려 #pragma omp 라는 구문을 해석하도록 한다.

$ man gcc
       -fopenmp
           Enable handling of OpenMP directives "#pragma omp" in C/C++ and
           "!$omp" in Fortran.  When -fopenmp is specified, the compiler
           generates parallel code according to the OpenMP Application Program
           Interface v3.0 <http://www.openmp.org/>.  This option implies
           -pthread, and thus is only supported on targets that have support
           for -pthread. 

[링크 : http://goulassoup.wordpress.com/2011/10/28/openmp-tutorial/]
[링크 : http://assaart.co.cc/wordpress/?p=59



+
우분투에서는 아래의 명령어로 openMP를 설치할 수 있다. gomp는 GNU OpenMP 의 약자이다
$ sudo apt-get install libgomp1 

'Programming > openMP' 카테고리의 다른 글

openMP 문서들  (0) 2012.06.18
openmp for문 나누기  (0) 2012.06.18
libgomp 공식 사이트 / 문서  (0) 2012.06.10
우분투에서 openMP 예제  (0) 2012.06.09
openMP  (0) 2011.02.05
Posted by 구차니
Programming/openCL & CUDA2012. 6. 7. 21:56
memcpy()와 비슷하게 dst, src 순서로 주소를 넣어주면 된다.
하지만, 그래픽 카드 메모리(device memory)와 메모리(host memory)를 구분지어 줘야하기 때문에
복사할 메모리의 방향과 종류를 정해주어야 한다.

일반적인 cuda 프로그래밍의 순서인
host -> device
cuda 계산
device -> host를 하기 위해서는

아래와 같이 한번씩 번갈아 해주면 될 듯?
cudaMemcpy(dev_memhost_mem, cudaMemcpyHostToDevice);
kernel_name<<< ... >>>(...);
cudaMemcpy(host_memdev_mem, cudaMemcpyDeviceToHost); 

5.8.2.18 cudaError_t cudaMemcpy (void *dst, const void *src, size_t count, enum cudaMemcpyKind kind)
Copies count bytes from the memory area pointed to by src to the memory area pointed to by dst, where kind is one of cudaMemcpyHostToHost, cudaMemcpyHostToDevice, cudaMemcpyDeviceToHost, or cudaMemcpyDevice-ToDevice, and specifies the direction of the copy. The memory areas may not overlap. Calling cudaMemcpy() with dst and src pointers that do not match the direction of the copy results in an undefined behavior.

Parameters:
dst - Destination memory address
src - Source memory address
count - Size in bytes to copy
kind - Type of transfer

Returns:
cudaSuccess, cudaErrorInvalidValue, cudaErrorInvalidDevicePointer, cudaErrorInvalidMemcpyDirection

5.28.3.9 enum cudaMemcpyKind
CUDA memory copy types

Enumerator:
cudaMemcpyHostToHost Host -> Host
cudaMemcpyHostToDevice Host -> Device
cudaMemcpyDeviceToHost Device -> Host
cudaMemcpyDeviceToDevice Device -> Device
cudaMemcpyDefault Default based unified virtual address space 

---
2012.07.11 추가
다시보니 cudaMemcpy(dst, src, direction); 의 양식이다.
다르게 보면 cudaMemcpy(To, From, dir_FromTo);
Posted by 구차니
Programming/openCL & CUDA2012. 6. 6. 08:18
예제파일 따라한다고 512 * 65535 개의 쓰레드를 계산하게 하는 int형 배열을 무려 3개나 할당!
4byte * 512 * 65535 = 대략 128MB?

아무튼 이걸 3개를 할당하니 378MB ..
근데 ION에다가 256MB만 할당해 놓은 시스템에서 저걸 돌리니
정상처럼 돌아가는데 결과는 전부 쓰레기값(전부 결과가 0이 나옴 -_-)
그리고 미친척(!) 7번 정도 실행하니 X윈도우까지 맛이 가서 ssh로 재시작 시키게 하는 센스 OTL



결론 : 메모리 사용량은 확실히 계산하고 malloc 해주자! 

'Programming > openCL & CUDA' 카테고리의 다른 글

cudaMalloc 시작 위치?  (0) 2012.07.11
cudemMemcpy()  (0) 2012.06.07
nvcc 는 int main()을 좋아해  (0) 2012.06.05
cuda 메모리별 접근 소요 클럭 사이클  (0) 2012.06.05
ubuntu 에서 vectorAdd 직접 컴파일 하기  (0) 2012.06.03
Posted by 구차니
Programming/openCL & CUDA2012. 6. 5. 22:39
심심(?)해서 void main()으로 해봤더니 요따구 에러 발생 -_-

vectorAdd.cu(11): warning: return type of function "main" must be "int"
vectorAdd.cu(11): warning: return type of function "main" must be "int"
vectorAdd.cu:11:11: error: ‘::main’ must return ‘int’ 

컴파일은 문제없다가 링킹에서 배째는 기분인데 -_-
Posted by 구차니
Programming/openCL & CUDA2012. 6. 5. 07:38
Nvidia cuda 공식 문서를 찾아보니, 메모리별 소요 클럭에 대한 내용이 자세히 나온다.

local memory 나 device memory 와 같이 off-chip memory(그래픽 카드에 내장되지 않은)는 400~800 클럭에 접근하고
shared memory나 register와 같이 on-ship memory(그래픽 카드에 내장된)는 10~20 클럭 정도에 접근이 가능하다.
constant memory는 2.0 에서 부터는 1회에 대해서 400~800 클럭이고, caching이 되면서 10~20 클럭에 읽어올 수 있다.

다르게 말하면 매우 빈번하게(클럭 빈도로는 40회 이상?) 읽어오는 데이터의 경우에는
무조건 shared memory에 읽어와서 빠르게 읽는게 전체 실행 속도에 유리할 것으로 보인다.
(읽어와서 다시 저장을 해야 한다면 80회 이상으로 늘어날지도?)

5.2  Maximize Utilization 
5.2.3  Multiprocessor Level 

If all input operands are registers, latency is caused by register dependencies, i.e. some of the input operands are written by some previous instruction(s) whose execution has not completed yet. In the case of a back-to-back register dependency  (i.e. some input operand is written by the previous instruction), the latency is equal to the execution time of the previous instruction and the warp schedulers must schedule instructions for different warps during that time. Execution time varies  depending on the instruction, but it is typically about 22 clock cycles for devices of compute capability 1.x and 2.x and about 11 clock cycles for devices of compute capability 3.0, which translates to 6 warps for devices of compute capability 1.x and  22 warps for devices of compute capability 2.x and higher (still assuming that warps execute instructions with maximum throughput, otherwise fewer warps are needed).  For devices of compute capability 2.1 and higher, this is also assuming enough instruction-level parallelism so that schedulers are always able to issue pairs of instructions for each warp. 

 If some input operand resides in off-chip memory, the latency is much higher: 400 to 800 clock cycles. The number of warps required to keep the warp schedulers busy during such high latency periods depends on the kernel code and its degree of  instruction-level parallelism. In general, more warps are required if the ratio of the number of instructions with no off-chip memory operands (i.e. arithmetic instructions most of the time) to the number of instructions with off-chip memory  operands is low (this ratio is commonly called the arithmetic intensity of the program). If this ratio is 15, for example, then to hide latencies of about 600 clock cycles, about 10 warps are required for devices of compute capability 1.x and about  40 for devices of compute capability 2.x and higher (with the same assumptions as in the previous paragraph).

5.3
Maximize Memory Throughput
The first step in maximizing overall memory throughput for the application is to minimize data transfers with low bandwidth.
That means minimizing data transfers between the host and the device, as detailed in Section 5.3.1, since these have much lower bandwidth than data transfers between global memory and the device.
That also means minimizing data transfers between global memory and the device by maximizing use of on-chip memory: shared memory and caches (i.e. L1/L2 caches available on devices of compute capability 2.x and higher, texture cache and constant cache available on all devices).



[출처 : CUDA_C_Programming_Guide.pdf]

Posted by 구차니