cuda api的采样主要cudart提供了profiler的接口，定义在cuda_profiler_api.h文件中，下面的代码是一个例子。
参考 https://blog.csdn.net/weixin_43603658/article/details/130441136，可以看到使用cudaProfilerStart和cudaProfilerEnd可以查看程序中指定段的性能数据。
nsys在这个基础上可能更加丰富了采样数据的类型，如ptrace看osrt的数据。
#include <iostream>
#include <cuda_runtime.h>
#include <cuda_profiler_api.h>// 定义向量长度
const int N = 1024;// 核函数，实现简单的向量加法
__global__ void vectorAdd(const float* a, const float* b, float* c) {int tid = blockIdx.x * blockDim.x + threadIdx.x;if (tid < N) {c[tid] = a[tid] + b[tid];}
}int main() {float* a, * b, * c;float* d_a, * d_b, * d_c;// 在主机上分配内存a = new float[N];b = new float[N];c = new float[N];// 在设备上分配内存cudaMalloc((void**)&d_a, N * sizeof(float));cudaMalloc((void**)&d_b, N * sizeof(float));cudaMalloc((void**)&d_c, N * sizeof(float));// 初始化主机上的向量数据for (int i = 0; i < N; i++) {a[i] = i;b[i] = i * 2;}// 将数据从主机复制到设备cudaMemcpy(d_a, a, N * sizeof(float), cudaMemcpyHostToDevice);cudaMemcpy(d_b, b, N * sizeof(float), cudaMemcpyHostToDevice);// 启动CUDA profilercudaProfilerStart();// 定义线程块和线程数量int blockSize = 256;int numBlocks = (N + blockSize - 1) / blockSize;// 调用核函数vectorAdd<<<numBlocks, blockSize>>>(d_a, d_b, d_c);// 检查核函数执行是否出错cudaError_t err = cudaGetLastError();if (err!= cudaSuccess) {std::cerr << "CUDA Error: " << cudaGetErrorString(err) << std::endl;}// 停止CUDA profilercudaProfilerStop();// 将结果从设备复制回主机cudaMemcpy(c, d_c, N * sizeof(float), cudaMemcpyDeviceToHost);// 释放设备内存cudaFree(d_a);cudaFree(d_b);cudaFree(d_c);// 释放主机内存delete[] a;delete[] b;delete[] c;return 0;
}
