20160324

동시성 프로그램의 이해
OpenMP는 scheduling class를 제공한다. static, dynamic, guided, runtime 등이 있다.
-> static은 전체 task N을 쓰레드의 수 t로 나눈 N/t를 각 쓰레드에 할당한다. 컴파일러가 쓰레드에 할당할 task를 결정한다.
-> Dynamic은 동적으로 쓰레드에 task를 할당한다. Load imbalance가 있다면 dynamic scheduling을 사용하는 것이 좋다. Dynamic은 어떤 단점이 있나? Dynamic의 경우에는 work queue와 scheduling을 위한 코드 등의 오버헤드가 있다. 이러한 cost가 있지만 load imbalance가 있으면 dynamic을 쓰는 것이 좋다.
-> Guided는 dynamic과 비슷하지만, 큰 task로부터 시작한다는 차이가 있다. 처음에는 크게 작업을 할당하고, 뒤로 갈수록 fine-grained task management를 한다는 특징이 있다.

OpenMP의 parallel for는 원래 끝난 다음에 쓰레드가 끝나기를 기다린다(barrier). 하지만 nowait을 주면 기다리지 않고 진행할 수 있다. 언제 쓰일 수 있을까? 개별 쓰레드 내부에서는 dependency가 있으나, 다른 쓰레드끼리는 dependency가 없을 때.

– sections directive는 non-iterative parallel 쓰레드를 생성할 수 있다. section마다 한 개의 쓰레드가 생성된다.
OpenMP sections.PNG
single는 한 개의 쓰레드만 해당 코드를 실행하도록 강제한다.
master는 쓰레드 0만 실행하는 부분이다.
critical은 한 번에 한 개의 쓰레드만 실행할 수 있는 코드이다. name 위치에 이름을 주면 lock처럼 사용할 수 있다.
ordered directive는 쓰레드들이 해당 구역의 코드를 순서대로 진행하도록 한다. 쓰레드가 병렬로 진행하지 못한다는 점에서 critical과 비슷하지만, 루프 순서로 진행해야 한다는 점에서 critical보다 강력한 제약이다.
OpenMP directives.PNG

– GPGPU는 general purpose gpu를 의미한다. GPU는 그래픽 처리를 위한 장치이다. GPU는 여러 개의 벡터 코어를 가지므로, 일반 목적에도 사용될 수 있다. CUDA는 Compute Unified Device Architecture의 약자.

– GPU는 자체로 DRAM을 갖는다. 많은 대역폭을 요구한다는 특징을 갖는다. GPU에서 사용하는 GDDR은 제공 대역폭이 크다는 특징이 있다. 연결 방식이 CPU에 연결된 메모리와 다르다. CPU에 연결된 메모리는 메모리 컨트롤러 채널에 제약된다. GPU에서는 메모리를 point-to-point connection으로 연결해서 대역폭이 매우 크다. GPU의 성능에 높은 메모리 대역폭은 매우 중요하다.

– CPU를 host, GPU를 device로 부른다. GPU에서 실행되는 코드를 kernel이라고 한다. HPC에서는 간단한 연산 루틴을 kernel이라고 한다. GPU는 내부에 DRAM을 가지며, device memory라고 한다. GPU에 데이터를 옮기는 것은 오버헤드가 있다. 따라서 너무 작은 문제는 그냥 CPU에서 실행하는 것이 좋을 수 있다. GPU 쓰레드는 매우 경량이다.

GPU는 1) multi-core, 2) multi-threaded, 3) vector-like 아키텍쳐이다. Streaming multiprocessor는 일종의 vector 프로세서이며, GPU는 여러 개의 SMP를 가지므로 multi-core 아키텍쳐이다. Intel 프로세서에서는 한 개의 물리적 코어를 두 개의 코어처럼 쓸 수 있다(하이퍼쓰레딩). 이를 위해 하이퍼쓰레딩에서는 context 정보를 저장해둔다. 그리고 실행 시점에 이를 불러와서 실행한다. GPU에서는 한 개의 SMP가 수백에서 수천개의 코어처럼 동작한다. 한 개의 물리적 코어가 아주 많은 논리적 코어를 생성할 수 있다. context 정보는 GPU에 모두 저장되며, context switch로 인한 오버헤드가 거의 없다. 따라서 소프트웨어 쓰레드는 하드웨어 쓰레드는 일대일 매핑된다. GPU의 SMP는 여덟 개의 streaming processor가 있으며, 각각은 같은 instruction을 갖고 다른 데이터에 적용한다. 한 개의 instruction으로 여덟 개의 데이터 요소를 연산할 수 있다.

– CUDA는 grid > thread block > thread의 계층을 갖는다. thread block 내에 thread를 갖게 된다. 같은 thread block 안에 있는 thread들은 1) shared memory를 가지며, 2) atomic operation을 지원하며, 3) barrier 동기화가 가능하다. thread block을 벗어나는 thread끼리는 이 세 가지가 지원되지 않는다. host가 kernel을 갖고 device에 grid를 생성한다. grid는 thread block으로 구성됨. thread block은 thread로 구성됨. 한편, grid 내의 thread block끼리는 global memory를 공유한다. 이 global memory는 CPU도 접근할 수 있다. Streaming processor 내에도 shared memory가 있기는 하지만, 외부에는 공유되지 않는다. cudaMalloc() 함수로 global memory에 메모리를 할당할 수 있다. cudaMemcpy() 함수는 host-host, host-device, device-host, device-device 사이의 메모리 복사를 지원한다. 비동기적으로 복사한다. double buffering을 사용한다. 한 개의 버퍼에서는 복사 작업을 수행하고, 다른 버퍼에서는 프로세싱 진행한다. 그리고 이 두 버퍼를 바꿔가며 진행. __device__ 함수는 재귀, static 변수, 가변 변수를 지원하지 않는다.

Advertisements
Tagged with: , ,
Posted in 1) Memo

Leave a Reply

Fill in your details below or click an icon to log in:

WordPress.com Logo

You are commenting using your WordPress.com account. Log Out /  Change )

Google+ photo

You are commenting using your Google+ account. Log Out /  Change )

Twitter picture

You are commenting using your Twitter account. Log Out /  Change )

Facebook photo

You are commenting using your Facebook account. Log Out /  Change )

w

Connecting to %s

Recent Posts
누적 방문자 수
  • 131,395 hits
%d bloggers like this: