메모리 계층 구조 - 실험 도서관

메모리 계층 구조는 컴퓨터 시스템의 성능 향상과 효율적인 자원 활용을 위해 다중 메모리 수준을 배치하고 조직하는 방법을 의미한다. GPU에서 메모리 계층 구조는 성능 최적화와 병렬 처리의 효율성을 극대화하기 위해 중요한 역할을 한다. 이 계층 구조는 일반적으로 다음과 같은 메모리 타입으로 구성된다:

레지스터 파일

레지스터 파일은 GPU 코어 내부에서 가장 빠르고 가장 작은 메모리이다. 레지스터는 매우 빠르지만 수가 제한적이므로, 계산 중간의 데이터를 일시적으로 저장한다.

로컬 메모리

로컬 메모리는 각 워프(Warp)나 스레드 블록이 독립적으로 사용하는 메모리 공간이다. 로컬 메모리는 주로 레지스터 초과 데이터를 저장하거나, 임시 데이터를 유지하는 데 사용된다.

공유 메모리

공유 메모리는 동일한 스레드 블록 내에서 모든 스레드가 접근 가능한 특수 고속 메모리이다. 공유 메모리는 동기화가 가능하며, 협력적인 데이터 공유와 통신에 효율적이다.

글로벌 메모리

글로벌 메모리는 GPU의 모든 스레드가 접근 가능한 가장 큰 메모리 공간이다. 그러나 접근 속도가 느리며, 따라서 캐시 등을 이용한 최적화가 필요하다.

텍스처 메모리

텍스처 메모리는 2D 또는 3D 이미지 데이터와 같은 읽기 전용 데이터를 저장하는 데 특화된 메모리이다. 텍스처 메모리는 캐시를 사용하여 자주 접근하는 데이터를 빠르게 사용할 수 있도록 설계되었다.

상수 메모리

상수 메모리는 읽기 전용 데이터를 저장하며, 스레드 간의 데이터 일관성을 보장한다. 모든 스레드가 동일한 상수 데이터를 자주 읽는 경우에 효율적이다. 캐시 메모리를 사용하여 성능을 최적화한다.

메모리 계층 구조의 특징

시간 지역성(Temporal Locality)

같은 데이터에 연속적으로 접근할 가능성을 나타낸다. 이를 통해 최근에 사용한 데이터가 다시 접근될 때 캐시 메모리를 활용해 성능 향상을 도모할 수 있다.

공간 지역성(Spatial Locality)

인접한 데이터가 연속적으로 접근될 가능성을 나타낸다. 메모리 데이터가 인접한 위치에 저장되면 캐시 라인을 읽어오는 방식으로 성능을 최적화할 수 있다.

메모리 계층 구조는 GPU에서 매우 중요한 역할을 하며, 각 계층은 특정한 성능상의 이유와 사용 사례를 염두에 두고 설계되었다.

메모리 계층 구조 최적화

메모리 통합(coalescing)

글로벌 메모리 접근에서 성능을 극대화하기 위해 메모리 통합을 활용한다. 메모리 통합은 인접한 메모리를 접근하는 요청을 하나의 메모리 트랜잭션으로 병합하는 기술이다. 이는 메모리 대역폭을 효율적으로 사용하여 성능을 향상시킨다.

공유 메모리의 효율적 사용

공유 메모리는 빠르지만 용량이 제한적이므로, 이를 효율적으로 사용하는 것이 중요하다. 예를 들어, 데이터 재사용이 빈번한 경우 공유 메모리에 데이터를 로드하고, 이를 반복적으로 사용하는 방식으로 성능을 개선할 수 있다.

뱅크 충돌(Bank Conflict) 회피

NVIDIA GPU에서는 공유 메모리가 여러 뱅크로 나뉘어져 있다. 만약 여러 스레드가 동일한 뱅크에 동시에 접근하려 한다면 충돌이 발생하여 성능이 저하될 수 있다. 따라서 데이터가 균등하게 분포되도록 배열하여 뱅크 충돌을 최소화하는 것이 중요하다.

예제 코드

다음은 GPU의 메모리 계층 구조를 효율적으로 활용하기 위한 간단한 CUDA 코드를 예시로 보여준다:

__global__ void matrixMul(float* C, float* A, float* B, int N) {
    // 공유 메모리 선언
    __shared__ float As[BLOCK_SIZE][BLOCK_SIZE];
    __shared__ float Bs[BLOCK_SIZE][BLOCK_SIZE];

    int bx = blockIdx.x;
    int by = blockIdx.y;
    int tx = threadIdx.x;
    int ty = threadIdx.y;

    // 결과 변수 초기화
    float Cvalue = 0;

    int Row = by * BLOCK_SIZE + ty;
    int Col = bx * BLOCK_SIZE + tx;

    // 블록 단위로 계산
    for (int t = 0; t < (N + BLOCK_SIZE - 1) / BLOCK_SIZE; ++t) {
        if (Row < N && t * BLOCK_SIZE + tx < N)
            As[ty][tx] = A[Row * N + t * BLOCK_SIZE + tx];
        else
            As[ty][tx] = 0;

        if (t * BLOCK_SIZE + ty < N && Col < N)
            Bs[ty][tx] = B[(t * BLOCK_SIZE + ty) * N + Col];
        else
            Bs[ty][tx] = 0;

        __syncthreads();

        for (int k = 0; k < BLOCK_SIZE; ++k)
            Cvalue += As[ty][k] * Bs[k][tx];

        __syncthreads();
    }

    if (Row < N && Col < N)
        C[Row * N + Col] = Cvalue;
}

위의 코드에서 공유 메모리를 사용하여 매트릭스 곱셈을 효율적으로 수행한다. 각 블록의 스레드들은 데이터를 공유 메모리에 로드하고, 그 후에 계산을 수행하여 글로벌 메모리 접근을 최소화한다.

GPU에서 메모리 계층 구조를 이해하고 이를 최적화하는 전략은 고성능 병렬 프로그램을 작성하는 데 매우 중요하다. 효율적인 메모리 접근 패턴과 최적화는 병렬 컴퓨팅 환경에서의 성능에 큰 영향을 미친다.