PCIe 相关
这是一个大坑,早晚要来填。。。
PCI BAR : base address register
这是一个大坑,早晚要来填。。。
PCI BAR : base address register
一篇关于 GPU 虚拟化 的 survey 文章,发表于 ACM Computing Surveys
Background 里面一段比较有意思的话:
On the contrary, the design of conventional processors is optimized
for reducing the execution time of sequential code on each core, thus adding complexity
to each core at the cost of offering fewer cores in the processor package. Conventional
processors typically use sophisticated control logic and large cache memories to efficiently
deal with conditional branches, pipeline stalls, and poor data locality.
传统处理器的目标是尽可能加速每个核心的串行执行时间,因此每个核心有大量处理分支预测,流水线延迟,data locality 相关的资源。代价是核心数量较少
之前做过劫持 cuda runtime 动态链接库的事情,在这里记录一下:
cudaError_t cudaMalloc ( void** devPtr, size_t size )
{
if(!fp) {
fp = fopen("/home/leenldk/sc/race/ASC2021-RACE/mem.out", "w");
}
cudaError_t (*lcudaMalloc) (void**, size_t) = (cudaError_t (*) (void**, size_t))dlsym(RTLD_NEXT, "cudaMalloc");
cudaError_t ret = lcudaMalloc(devPtr, size);
printf("cudaMalloc size : %#lx %#lx\n", size, (size_t)(*devPtr));
fprintf(fp, "cudaMalloc size : %#lx %#lx\n", size, (size_t)(*devPtr));
fflush(fp);
return ret;
}
编译为动态链接库,使用 LD_PRELOAD
预加载
update 2022.11.2
C++ 动态库 加载,卸载 时调用函数 :
static void init() __attribute__((constructor));
void init() {}
static void fini() __attribute__((destructor));
void fini() {}
linux 将外设抽象为 /dev
下的文件,通过统一的文件读写接口访问外设
linux与外设通信:
managed memory : 使用 linux heterogeneous memory management (HMM), device 和 host 端可以以相同指针访问同一块内存
coherent memory : 可以在 kernel 运行时执行对 host 和其他 peer 可见的原子操作,通过不 cache 内存实现
non-coherent memory : device 端 cache 的内存,修改不实时可见
direct dispatch : runtime 直接将操作发送至 AQL 队列
device side malloc
HIP 支持两种 static lib: