딥러닝 모델에서 CUDA 를 활용한 연산을 수행할 때의 메모리 배치 전략, shared Memory vs Global Memory 전략

✅ 딥러닝 연산 대상별 적합한 메모리 전략

가중치 (W, b)	반복 사용됨, 크기 고정	✅ Global memory (고정), 필요한 경우 shared에 캐싱
입력값 (x)	배치 단위로 바뀜	✅ Global memory (입력마다 갱신)
출력값 (z, a)	다음 레이어 입력이 됨	✅ Global memory (혹은 register로 재사용)
중간 연산 결과 (e.g. matmul 결과)	다음 연산에 바로 쓰임	🔄 Shared memory로 저장 시 이득
활성화 함수 등 커널 내 임시값	kernel 내부 계산용	✅ Register (가장 빠름)

 

딥러닝 프레임워크에서 레이어별 연산이 GPU 상에서 어떤 메모리를 어떻게 사용하는지를 이해하는 것은 성능 최적화의 핵심

 

✅ 레이어별 메모리 흐름도 요약

📘 표기

표기                     의미

GM	Global Memory (GPU 전역 메모리)
SM	Shared Memory (block 내 빠른 협업 메모리)
REG	Register (thread 내 연산용 임시 저장)

 

💡 팁: 메모리 흐름 최적화를 위한 규칙

Global ↔ Shared 복사 최소화	필요한 부분만 블록당 복사
가중치는 캐시 or SM에 보관	반복 사용 시 효율적
임시 계산 결과는 REG 활용	빠르고 thread-local
연산 순서 맞춰 메모리 흐름 조절	shared reuse 고려

 

✅ Shared Memory vs Global Memory: 핵심 비교표

위치	각 thread block 내부에 존재	GPU 전체에서 접근 가능
범위 (Scope)	블록 내의 쓰레드들만 공유	모든 쓰레드에서 접근 가능
속도	매우 빠름 (10~100배 빠름)	느림 (400~800 cycle 지연)
수명	커널 실행 중에만 존재	할당 후 명시적 해제 전까지 유지
용량	작음 (대개 48KB/block)	큼 (수 GB까지 가능)
사용 용도	블록 내 데이터 공유, 중간 계산 캐시	모델 파라미터, 입력/출력 저장 등
할당 방식	커널 내 __shared__ 선언	cudaMalloc, CuPy, PyTorch 등
메모리 정렬(coalescing)	정렬과 무관 (빠름)	정렬 필수 (그렇지 않으면 느림)

 

 

📌 어디에 어떻게 쓰면 되나?

모델의 파라미터, 입력/출력	Global memory (크고 공유 가능)
커널 내부에서 반복 사용하는 중간값	Shared memory (속도 매우 빠름)
한 block 내 연산 병렬화	Shared memory (스레드 협업)
전 모델 실행 내내 유지할 값	Global memory (Persistent)
임시 계산 값, sum 누적 등	Register or shared memory

 

---

✅ 요약: 비유로 정리

Global Memory	🏢 건물 내 창고 (크지만 느리고 복잡)
Shared Memory	🏠 작업팀만 쓰는 공구함 (작지만 매우 빠름)
Register	👷 개인 주머니 속 공구 (가장 빠름, 작음)

 

---

✅ 결론

고성능 CUDA 커널은 Global Memory에서 필요한 데이터만 가져와서 Shared Memory에 캐시하고,
그 후에 Register를 활용해 빠르게 계산한 뒤, 다시 결과를 Global에 쓰는 방식이 이상적입니다.