분류 전체보기62 dynamic parallelism을 활용한 quick sort Dynamic Parallelism의 간단한 개념 설명) https://hayunjong83.tistory.com/22 CUDA 샘플 코드에서는 dynamic parallelism을 설명하기 위해, quick sort 코드를 제공하고 있다. depth가 너무 커지거나 일정 임계값 이하의 원소를 정렬할 때는 선택정렬(selection sort)를 사용한다. 퀵 정렬 알고리즘 자체는 기존의 흔한 구현과 다르지 않다. 피벗을 기준으로 왼쪽과 오른쪽을 다시 정렬할 스레드를 선언하여 재귀적으로 다시 커널을 호출한다. 이 때 dynamic parallelism을 이용하므로, 중간 결과를 다시 호스트로 memcpy할 필요가 없으며, stream을 이용하므로 각 부분의 정렬이 다른 부분의 정렬과의 무의미한 동기화를 위.. 2020. 3. 6. Dynamic Parallelism 이란? 케플러 이전의 GPU에서는 오직 CPU만이 디바이스에 작업을 할당할 수 있었다. 하지만 dynamic parallelism이 도입되면서 GPU의 작업 중인 스레드가 새로운 그리드를 생성하고 작업을 할당할 수 있게 되었다. 실행 중인 CUDA 커널은 새롭게 중첩된(nested) 작업을 생성하고 동기화 할 수 있다. 이 때 원래의 커널을 부모 커널(Parent CUDA Kernel)이라고 하고, 새롭게 생성된 커널을 자식 커널(Child CUDA Kernel)이라고 한다. 부모 커널은 자식 커널에 의해 생성된 결과를 사용할 수 있으며, 이 때 CPU가 개입될 필요가 없다. 또한 재귀 호출도 가능하다. dynamic parallelism을 이용한다면, 한 커널의 결과를 입력으로 하는 새로운 커널을 시작할 때 .. 2020. 3. 5. CUDA Event를 이용한 커널(SAXPY) 실행시간 측정 호스트의 타이머 함수를 사용할 경우, cudaDeviceSynchronize( )와 같은 명시적인 배리어 함수를 사용해야한다. 이 경우 GPU 파이프라인이 stall되는 문제 등이 발생될 수 있다. 따라서 CUDA 런타임에서 제공하는 event API를 사용해서 커널의 실행 시간을 측정하는 방식을 이용하는 것이 좋다. CUDA 이벤트 데이터 타입인 cudaEvent_t 를 사용하는 다음의 함수들을 사용한다. __host__ cudaError_t cudaEventCreate( cudaEvent_t* event) __host__ __device__ cudaError_t cudaEventDestroy(cudaEvent_t* event) __host__ __device__ cudaError_t cudaEvent.. 2020. 3. 4. SAXPY(Single-precision A-X Plus Y) CUDA 프로그래밍 예제로 가장 흔히 사용하는 SAXPY(Single-precision A-X Plus Y)는 스칼라 곱과 백터 합이 결합된 함수로서 GPU 병렬연산의 기본적인 내용을 설명하기에 적합하다. 이번 포스트에서는 C++를 이용한 단순한 SAXPY 코드와 CUBLAS 라이브러리를 이용한 SAXPY 실행방법을 살펴본다. 단순하게 쓰레드의 인덱스를 결정하기 위해서, 그리드와 블록 모두 1차원 형태로 가정하였다. restrict 포인터는 메모리 접근을 최적화하기 위해 사용하였다. cublas 라이브러를 이용하는 경우, cublasSaxpy( )를 이용해서 동일한 연산을 시행할 수 있다. 위와 비교해봤을 때, 직접 SAXPY에 대한 커널 작성이 필요없음을 알 수 있다. 참고로 cublasAlloc( ).. 2020. 3. 3. 이전 1 ··· 10 11 12 13 14 15 16 다음
