fake_gpu_utilization

fake_gpu_utilization.cu

#include <string>
#include <unistd.h>
#include <iostream>
#include <nvml.h>

#define cudaSafeCall(call) \
do { \
    cudaError_t err = call; \
    if (cudaSuccess != err) { \
        fprintf (stderr, "Cuda error in file '%s' in line %i : %s.", \
                 __FILE__, __LINE__, cudaGetErrorString(err) ); \
         exit(EXIT_FAILURE); \
    } \
} while (0)

__global__ void one_busy_thread(bool* done) {
  if (threadIdx.x == 0) {
    while (!done[0]) {
    }
  }
}

int main(int argc, char** argv) {

    nvmlInit();

    bool* d_done = nullptr;
    constexpr int size = sizeof(bool);

    int device_id = argc == 1 ? 0 : std::stoi(argv[1]);

    cudaSafeCall(cudaSetDevice(device_id));
    cudaSafeCall(cudaMalloc(&d_done, size));
    cudaSafeCall(cudaMemset(d_done, 0, size));

    dim3 grid = 1;
    dim3 block = 1;

    one_busy_thread<<<grid, block>>>(d_done);

    cudaError_t kernel_error = cudaGetLastError();
    if (kernel_error != cudaSuccess) {
         std::cout << "Error launching kernel: " << cudaGetErrorString(kernel_error) << std::endl;
    }

    sleep(2);
    unsigned int power_limit, power_usage;
    nvmlUtilization_t utilization;
    nvmlDevice_t device;
    nvmlDeviceGetHandleByIndex_v2(device_id, &device);

    while (true) {
        nvmlReturn_t ret = nvmlDeviceGetUtilizationRates(device, &utilization);
        nvmlDeviceGetEnforcedPowerLimit(device, &power_limit);
        nvmlDeviceGetPowerUsage(device, &power_usage);
        float power_usage_pct = 100.0 * power_usage / power_limit;
        if (ret != NVML_SUCCESS) {
                std::cout << "NVML error: " << nvmlErrorString(ret) << std::endl;
                nvmlShutdown();
                return -1;
        }
        std::cout << "Utilization when running with " << grid.x << " block of " << block.x << " thread(s): " << utilization.gpu << "%" << ", power usage: " << power_usage_pct << "%" << std::endl;
        sleep(1);
    }

    return 0;
}

To build and run:

nvcc utilization_example.cu -o utilization_example -l nvidia-ml -arch=sm_75 && ./utilization_example

Output:

Utilization when running with 1 block of 1 thread(s): 100%, power usage: 29.5043%
Utilization when running with 1 block of 1 thread(s): 100%, power usage: 29.6125%
Utilization when running with 1 block of 1 thread(s): 100%, power usage: 29.6718%
Utilization when running with 1 block of 1 thread(s): 100%, power usage: 29.6639%
Utilization when running with 1 block of 1 thread(s): 100%, power usage: 29.7114%
Utilization when running with 1 block of 1 thread(s): 100%, power usage: 29.8229%
Utilization when running with 1 block of 1 thread(s): 100%, power usage: 29.8104%

A better metric of utilization would be achieved occupancy, which in the old CUPTI metric API was called achieved_occupancy, and in the new PerfWorks API is called sm__warps_active.avg.pct_of_peak_sustained_active

See the table here