04. Digging into NVIDIA kernel module

Open GPU kernel module repository

https://github.com/NVIDIA/open-gpu-kernel-modules

Kernel module 구성

위 repository를 clone해서 빌드하면 아래 다섯 개의 .ko 파일이 생성됩니다.

• nvidia.ko : X server, OpenGL, OpenCL, CUDA와 같은 user space library에서 공통적으로 사용하는 GPU 제어 인터페이스
• nvidia-drm.ko: Direct Rendering Manager (디스플레이 및 렌더링 인터페이스)
• nvidia-uvm.ko : CUDA Unified Memory에 필요한 CPU-GPU간 memory migration을 제어하는 인터페이스
• nvidia-modeset.ko: GPU의 display engine을 제어하는 데 사용되는 communication interface
• nvidia-peermem.ko: GPUDirect RDMA 인터페이스

ftrace를 사용해서 호출되는 kernel module 함수 추적해보기

ftrace는 kernel에서 호출되는 함수나 이벤트를 추적해볼 수 있게 해주는 linux tool입니다.
CUDA application에서 Host to Device memory 복사를 수행하면 kernel module에서는 어떤 함수가 호출되는지 살펴봤습니다.
Application 소스 코드는 아래와 같습니다.

int main(void)
{
  int *h, *d;
  h = (int *)malloc(sizeof(int));
  cudaMalloc(&d, sizeof(int));

  cudaMemcpy(d, h, sizeof(int), cudaMemcpyHostToDevice);

  free(h);
  cudaFree(d);

  return 0;
}

아래 command를 실행해서 ftrace를 설정하고 실행합니다.

su
# Move to ftrace-related directory. This should be accessed as a superuser. 
cd /sys/kernel/debug/tracing
# Set tracer mode to function_graph.
echo function_graph > current_tracer
# Set to collect only traces from nvidia.ko and nvidia_uvm.ko modules 
echo ':mod:nvidia' > set_trace_filter
echo ':mod:nvidia_uvm' >> set_trace_filter
# Start tracing 
echo 1 > tracing_on
### Run CUDA application ###
# Stop tracing 
echo 0 > tracing_on

trace 파일에 적힌 내용을 확인합니다. 내용이 너무 많아서 핵심적인 부분만 clip 했습니다.
nv_alloc_pages 함수가 memory 할당에 관여하는 것으로 보입니다.

# tracer: function_graph
#
# CPU  DURATION                  FUNCTION CALLS
# |     |   |                     |   |   |   |
... skipped ...
  67)               |        nv_alloc_pages [nvidia]() {
  67)   0.240 us    |          nv_printf [nvidia]();
  67)   0.218 us    |          nv_printf [nvidia]();
  67)   0.268 us    |          nv_encode_caching [nvidia]();
  67)               |          nvos_create_alloc [nvidia]() {
  67)   0.663 us    |            os_alloc_mem [nvidia]();
  67)   0.219 us    |            os_get_current_process [nvidia]();
  67)   4.582 us    |          }
  67)               |          nv_alloc_contig_pages [nvidia]() {
  67)   0.218 us    |            nv_printf [nvidia]();
  67)   0.226 us    |            os_is_xen_dom0 [nvidia]();
  67)               |            nv_compute_gfp_mask [nvidia]() {
  67)   0.282 us    |              nv_requires_dma_remap [nvidia]();
  67)   1.081 us    |            }
  67)   0.382 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.166 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.242 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.157 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.151 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.153 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.151 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.153 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.150 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.149 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.145 us    |            nv_get_kern_phys_address [nvidia]();
  67)   0.149 us    |            nv_get_kern_phys_address [nvidia]();
  67) ! 370.421 us  |          }
... skipped ...

kernel module 소스 코드에서 nv_alloc_pages 함수를 검색합니다.
해당하는 소스 파일을 열어서 함수의 선언과 정의를 확인할 수 있습니다.

dtb05045@a1:~/open-gpu-kernel-modules$ grep -r 'nv_alloc_pages'
kernel-open/common/inc/nv.h:NV_STATUS  NV_API_CALL  nv_alloc_pages           (nv_state_t *, NvU32, NvBool, NvU32, NvBool, NvBool, NvU64 *, void **);
kernel-open/nvidia/nv.c:NV_STATUS NV_API_CALL nv_alloc_pages(
kernel-open/nvidia/nv.c:    nv_printf(NV_DBG_MEMINFO, "NVRM: VM: nv_alloc_pages: %d pages\n", page_count);
src/nvidia/arch/nvalloc/unix/src/os.c:        status = nv_alloc_pages(
src/nvidia/arch/nvalloc/unix/include/nv.h:NV_STATUS  NV_API_CALL  nv_alloc_pages           (nv_state_t *, NvU32, NvBool, NvU32, NvBool, NvBool, NvU64 *, void **);
tags:nv_alloc_pages     kernel-open/nvidia/nv.c /^NV_STATUS NV_API_CALL nv_alloc_pages($/;"     f