1.8086模拟器8086tiny源码分析(5)执行mov指令(二)
2.MLIR Operation 与 Op
3.3d稀疏卷积——spconv源码剖析（三）
4.Linux内核中的源码open方法

8086模拟器8086tiny源码分析(5)执行mov指令(二)

       本文继续解析tiny模拟器中的MO指令，集中于MOV reg,源码 r/m指令的实现。首先，源码通过xlat_opcode_id赋值为9，源码额外参数extra设置为8，源码为后续解析打下基础。源码生成 GIF 源码核心部分在于理解OP(=)的源码操作，其完成的源码是寄存器与内存或另一个寄存器间的数据移动。

       进一步分析，源码MEM_OP和R_M_OP分别对应内存操作与寄存器与内存间的源码拷贝，前者是源码基本内存操作，后者完成具体数据移动任务。源码而op_to_addr和op_from_addr则是源码关键变量，前者代表目的源码位置，后者代表源位置。源码具体赋值依赖于是否需要解码mod、rm、reg三个指令字段。spring源码模板下载

       当i_mod_size为真时，解码这三个字段，并结合d和w字段，确定操作数。这由DECODE_RM_REG宏完成。在这里，op_to_addr是目的位置（寄存器或内存），op_from_addr是源位置。指令数据移动方向的关键在于i_d变量。如果该变量为真，则表示源操作数与目的操作数需进行交换。

       至此，对MOV reg, r/m指令的解析告一段落。通过明确指令字段、操作变量的赋值与交换条件，tiny模拟器成功实现这一重要指令的执行，为深入理解架构与模拟器设计提供了基础。彩票双语源码

MLIR Operation 与 Op

       MLIR的核心组件之一是Operation，它是Dialect中的基本语义单元，就像方言中的词汇一样，构成了整个代码表示的基石。深入理解MLIR的运作，关键在于掌握Operation类和Op类的运作机制。

       Operation类是MLIR中的核心数据结构，它在 mlir/include/mlir/IR/Operation.h 中被定义，为操作提供了丰富的接口，包括静态create函数，用于创建不同类型的Operation实例。例如，当我们创建一个FuncOp时，会先通过Operation::create调用，再通过cast转换为具体类。此外，Operation还负责操作数、买源码是啥结果、属性和blocks的管理，以及操作的遍历等。

       Op类则是所有具体Operation的基类，它在mlir/include/mlir/IR/OpBase.td中定义，这个文件实际上是TableGen语言编写的。在定义新的Dialect时，开发者需要在Ops.td文件中引入OpBase.td，基于这个基类扩展出符合特定Dialect需求的Op类。

       整个过程涉及到TableGen工具的编译和链接，将这些定义转化为编译时可用的代码。通过理解并操作这些基础类，开发者可以构建和优化复杂的编译逻辑。更多关于Op类的具体细节和功能，还需进一步研究mlir的源码和官方文档。

3d稀疏卷积——spconv源码剖析（三）

       构建Rulebook

       下面看ops.get_indice_pairs，位于：spconv/ops.py

       构建Rulebook由ops.get_indice_pairs接口完成

       get_indice_pairs函数具体实现：

       主要就是梦幻铅笔源码在哪完成了一些参数的校验和预处理。首先，对于3d普通稀疏卷积，根据输入shape大小，kernel size，stride等参数计算出输出输出shape，子流行稀疏卷积就不必计算了，输出shape和输入shape一样大小

       准备好参数之后就进入最核心的get_indice_pairs函数。因为spconv通过torch.ops.load_library加载.so文件注册,所以这里通torch.ops.spconv.get_indice_pairs这种方式来调用该函数。

       算子注册：在src/spconv/all.cc文件中通过Pytorch提供的OP Register(算子注册的方式)对底层c++ api进行了注册，可以python接口形式调用c++算子

       同C++ extension方式一样，OP Register也是Pytorch提供的一种底层扩展算子注册的方式。注册的算子可以通过 torch.xxx或者 tensor.xxx的方式进行调用，该方式同样与pytorch源码解耦，增加和修改算子不需要重新编译pytorch源码。用该方式注册一个新的算子，流程非常简单：先编写C++相关的算子实现，然后通过pytorch底层的注册接口（torch::RegisterOperators），将该算子注册即可。

       构建Rulebook实际通过python接口get_indice_pairs调用src/spconv/spconv_ops.cc文件种的getIndicePairs函数

       代码位于：src/spconv/spconv_ops.cc

       分析getIndicePairs直接将重心锁定在GPU逻辑部分，并且子流行3d稀疏卷积和正常3d稀疏卷积分开讨论，优先子流行3d稀疏卷积。

       代码中最重要的3个变量分别为：indicePairs，indiceNum和gridOut，其建立过程如下：

       indicePairs代表了稀疏卷积输入输出的映射规则，即Input Hash Table 和 Output Hash Table。这里分配理论最大的内存，它的shape为{ 2,kernelVolume,numAct}，2表示输入和输出两个方向，kernelVolume为卷积核的volume size。例如一个3x3x3的卷积核，其volume size就是(3*3*3)。numAct表示输入有效(active)特征的数量。indiceNum用于保存卷积核每一个位置上的总的计算的次数，indiceNum对应中的count

       代码中关于gpu建立rulebook调用create_submconv_indice_pair_cuda函数来完成，下面具体分析下create_submconv_indice_pair_cuda函数

       子流线稀疏卷积

       子流线稀疏卷积是调用create_submconv_indice_pair_cuda函数来构建rulebook

       在create_submconv_indice_pair_cuda大可不必深究以下动态分发机制的运行原理。

       直接将重心锁定在核函数：

       prepareSubMGridKernel核函数中grid_size和block_size实则都是用的整形变量。其中block_size为tv::cuda::CUDA_NUM_THREADS,在include/tensorview/cuda_utils.h文件中定义，大小为。而grid_size大小通过tv::cuda::getBlocks(numActIn)计算得到,其中numActIn表示有效(active)输入数据的数量。

       prepareSubMGridKernel作用：建立输出张量坐标(通过index表示)到输出序号之间的一张哈希表

       见：include/spconv/indice.cu.h

       这里计算index换了一种模板加递归的写法，看起来比较复杂而已。令：new_indicesIn = indicesIn.data()，可以推导得出index为：

       ArrayIndexRowMajor位于include/tensorview/tensorview.h，其递归调用写法如下：

       接着看核函数getSubMIndicePairsKernel3：

       位于：include/spconv/indice.cu.h

       看：

       上述写法类似我们函数中常见的循环的写法，具体可以查看include/tensorview/kernel_utils.h

       NumILP按默认值等于1的话，其stride也是gridDim.x*blockDim.x。索引最大值要小于该线程块的线程上限索引blockDim.x * gridDim.x，功能与下面代码类似：

       参考： blog.csdn.net/ChuiGeDaQ...

Linux内核中的open方法

       在Linux系统中，使用open方法打开文件是一个基本操作，本文将从源码角度解析Linux内核中的open方法是如何实现文件打开功能的。

       在Linux内核中，open方法的实现主要涉及几个关键步骤。首先，调用do_sys_open方法作为主要入口。该方法内进行了一系列初始化操作，主要包括：

       1. 调用build_open_flags方法，初始化struct open_flags实例op。

       2. 调用getname方法，分配并初始化struct filename实例tmp。

       3. 调用get_unused_fd_flags方法获取一个未被使用的文件描述符fd。

       4. 调用do_filp_open方法，继续执行open操作，并将返回值赋值给类型为struct file的实例指针f。

       5. 如果do_filp_open成功，则调用fd_install方法，建立从fd到struct file的对应关系。

       6. 最后，返回fd给用户。

       在do_filp_open方法中，进一步执行open操作。主要步骤包括：

       1. 调用set_nameidata方法，初始化struct nameidata类型实例nd。

       2. 调用path_openat方法继续执行open操作。

       3. 在path_openat方法内，首先调用alloc_empty_file方法分配一个空的struct file实例。

       4. 然后调用path_init、link_path_walk、do_last等方法执行后续的open操作。

       5. 其中，path_init方法初始化struct nameidata实例中的path、root、inode等字段。

       6. link_path_walk方法处理路径组件，最终使struct nameidata实例指向目标路径。

       7. 最后，do_last方法完成剩余的open操作，如查找路径中的最后一个组件，执行open逻辑。

       整个open过程涉及多层函数调用，从系统调用到内核源码的详细实现，体现了Linux内核设计的复杂性和层次性。通过逐步解析这些步骤，可以更直观地理解Linux内核如何实现文件打开功能。