HMM support in UVM

NVIDIA Open GPU Kernel Modules Version

515.57 Release

Does this happen with the proprietary driver (of the same version) as well?

Yes

Operating System and Version

Ubuntu 20.04.1 LTS

Kernel Release

5.13.0-1029-aws

Hardware: GPU

NVIDIA T4

Describe the bug

Hello!

HMM support has been mentioned in several NVIDIA docs and presentations since 2017 (including the announcement of the open-source kernel modules), but it seems to be disabled here (and doesn't work when using the proprietary driver).

https://github.com/NVIDIA/open-gpu-kernel-modules/blob/d8f3bcff924776518f1e63286537c3cf365289ac/kernel-open/nvidia-uvm/uvm_hmm.h#L35-L50

I assume the referenced bug/task is internal. Is there any information you can share on what additional work needs to happen to enable UVM-HMM (or potentially a timeline?).

See references and a repro below.

To Reproduce

I'm testing HMM using the following code (from one of the presentations linked below):

#include <stdio.h>

#define LEN sizeof(int)

__global__ void
compute_this(int *pDataFromCpu)
{
    atomicAdd(pDataFromCpu, 1);
}

int main(void)
{
    int *pData = (int*)malloc(LEN);
    *pData = 0;

    // Run on GPU:
    compute_this<<<512,1000>>>(pData);

    cudaDeviceSynchronize();

    printf("Results: %d\n", *pData);

    free(pData);
    return 0;
}

It currently just prints Results: 0 and the following message is in the output of dmesg

[ 3896.223804] NVRM: Xid (PCI:0000:00:1e): 31, pid=42009, name=a.out, Ch 00000007, intr 00000000. MMU Fault: ENGINE GRAPHICS GPCCLIENT_T1_0 faulted @ 0x5558_a1468000. Fault is of type FAULT_PDE ACCESS_TYPE_VIRT_ATOMIC

Bug Incidence

Always

nvidia-bug-report.log.gz

N/A

More Info

References:

• 2017 GTC talk - Using HMM to Blur the Lines between CPU and GPU Programming
• 2019 GTC talk - Memory Management on Modern GPU Architectures
• The announcement post for this repo
• A 2017 Red Hat summit presentation - GPUs: HMM: Heterogeneous Memory
  Management

Start with kernel-open/nvidia-uvm/uvm_hmm.h at the referenced lines and review the linked HMM presentations. Run the supplied reproducer on the stated Ubuntu, kernel, and NVIDIA T4 setup, then inspect the reported dmesg MMU fault. Done would require a documented implementation path or enabled UVM-HMM support that avoids the reported failure.