CUDA Event Timing과 Benchmark
최적화의 첫 단계는 “빠른 것 같다”가 아니라 “같은 구간을 반복해서 재면 더 빠르다”이다.
host setup H2D copy kernel D2H copy verify
무엇을 잴 것인가
CPU timer로 전체 시간을 재면 host setup, memory copy, synchronization이 섞인다. kernel 자체 시간을 보려면 보통 CUDA event를 같은 stream에 기록한다.
cudaEventRecord(start);
kernel<<<grid, block>>>(...);
cudaEventRecord(stop);
cudaEventSynchronize(stop);
cudaEventElapsedTime(&ms, start, stop);
Colab에서의 현실
Colab에서는 Nsight Compute hardware counter 접근이 제한될 수 있다. 그래도 CUDA event, PyTorch torch.utils.benchmark, warmup loop만으로도 학습용 비교는 충분히 가능하다.
확인
- CPU timer와 CUDA event가 재는 구간은 어떻게 다른가?
- warmup을 하지 않으면 어떤 값이 섞일 수 있는가?
- 최적화 전후 비교에서 input shape를 고정해야 하는 이유는 무엇인가?