Rdma data transmission method, network device, system and electronic device

Abstract

The present disclosure provides an RDMA data transmission method and network device and system, relates to the technical field of communication, in particular to the technical field of chip processor, computing power cluster, collective communication operation, large model and the like. The method is applied to a network device comprising an xPU and an RNIC, the xPU comprising a first engine, the RNIC comprising a second engine in communication with the first engine, a WQE Buffer and an RDMA engine in communication with the second engine. The method comprises: the first engine assembles a WE based on a hardware offloading asynchronous copy instruction set and transmits the WE to the second engine; the second engine stores the WE in the WQE Buffer and transmits the WE to the RDMA engine; the RDMA engine performs data processing based on the WE, the data processing comprising data transmission, memory access and queue management; and the first engine receives feedback of the data processing transmitted via the second engine. The RDMA data transmission task is directly offloaded to a dedicated engine through a hardware asynchronous instruction set, the RDMA data transmission task is decoupled from the computing resources of the system, and an efficient and low-latency RDMA data transmission solution is provided for data transmission intensive scenarios such as MoE models without relying on CPU / GPU resources.

Description

TECHNICAL FIELD

[0001] The present disclosure relates to the technical field of communication, in particular to the technical field of chip processor, computing power cluster, set communication operation, large model, and the like, and more particularly to an RDMA data transmission method, a network device, a system, an electronic device, a computer readable storage medium, and a computer program product. BACKGROUND

[0002] In a generative large language model, a MoE (Mixture of Experts) model can process different tasks through multiple "expert" networks. In the MoE scenario, there is a large amount of scattered small data block transmission demand.

[0003] RDMA (Remote Direct Memory Access) protocol is a network protocol that allows a computer to directly access the memory of a remote computer, commonly used in HPC (High Performance Computing), data centers, storage networks, and the like. SUMMARY

[0004] Embodiments of the present disclosure provide an RDMA data transmission method, a network device, a system, an electronic device, a computer readable storage medium, and a computer program product.

[0005] In a first aspect, embodiments of the present disclosure provide an RDMA data transmission method, applied to a network device including a processor xPU and an RDMA network interface controller RNIC, the xPU including a first engine, the RNIC including a second engine in communication with the first engine, a work queue buffer WQE Buffer, and an RDMA engine in communication with the second engine, the method including: the first engine assembling a work element WE based on a hardware offload asynchronous copy instruction set, and transmitting the WE to the second engine; the second engine storing the WE in the WQE Buffer, and transmitting the WE to the RDMA engine; the RDMA engine performing data processing based on the WE, the data processing including data transmission, memory access, and queue management; and the first engine receiving feedback of the data processing transmitted via the second engine.

[0006] In a second aspect, embodiments of the present disclosure provide another method for RDMA data transfer, which is applied to a network system including a plurality of network devices, each of the network devices including a processor xPU and a RDMA network interface controller RNIC, the xPU including a first engine, the RNIC including a second engine in communication with the first engine, a work queue buffer WQE Buffer, and a RDMA engine in communication with the second engine, the method including: assembling, by the first engine of a first network device, a work element WE based on a hardware offloaded asynchronous copy instruction set, and transmitting the WE to the second engine of the first network device; storing, by the second engine of the first network device, the WE in the WQE Buffer, and transmitting the WE to the RDMA engine of the first network device; initiating, by the RDMA engine of the first network device, a request to a second network device based on the WE; and completing, by the second network device, the request, and feeding back to the first engine of the first network device, wherein the first network device and the second network device are different network devices in the plurality of network devices.

[0007] In a third aspect, embodiments of the present disclosure provide a network device for RDMA data transfer, the network device including: a processor xPU including a first engine; and a RDMA network interface controller RNIC including a second engine in communication with the first engine, a work queue buffer WQE Buffer, and a RDMA engine in communication with the second engine, wherein the first engine is configured to assemble a work element WE based on a hardware offloaded asynchronous copy instruction set, and transmit the WE to the second engine; the second engine is configured to store the WE in the WQE Buffer, and transmit the WE to the RDMA engine; and the RDMA engine is configured to perform data processing based on the WE, and transmit feedback of the data processing to the first engine via the second engine, the data processing including data transfer, memory access, and queue management.

[0008] In a fourth aspect, embodiments of the present disclosure provide an RDMA data transmission network system, the network system comprising: a plurality of network devices, a network device comprising a processor xPU and a RDMA network interface controller RNIC, the xPU comprising a first engine, the RNIC comprising a second engine in communication with the first engine, a work queue buffer WQE Buffer, and a RDMA engine in communication with the second engine, wherein the first engine of the first network device is configured to assemble a work element WE based on a hardware offloaded asynchronous copy instruction set and transmit the WE to the second engine of the first network device; the second engine of the first network device is configured to store the WE in the WQE Buffer and transmit to the RDMA engine of the first network device; the RDMA engine of the first network device is configured to initiate a request to a second network device based on the WE; the second network device is configured to complete the request and feedback to the first engine of the first network device, wherein the first network device and the second network device are different network devices in the plurality of network devices.

[0009] In a fifth aspect, embodiments of the present disclosure provide an electronic device, the electronic device comprising: at least one processor; and a memory communicatively connected with the at least one processor; wherein the memory stores instructions executable by the at least one processor, the instructions being executed by the at least one processor to enable the at least one processor to implement the RDMA data transmission method as described in the first aspect and the second aspect.

[0010] In a sixth aspect, embodiments of the present disclosure provide a non-transitory computer-readable storage medium storing computer instructions for causing a computer to implement the RDMA data transmission method as described in the first aspect and the second aspect.

[0011] In a seventh aspect, embodiments of the present disclosure provide a computer program product comprising a computer program which, when executed by a processor, is capable of implementing the RDMA data transmission method as described in the first aspect and the second aspect.

[0012] In the RDMA data transfer scheme provided in this disclosure, the first engine is directly integrated inside the xPU, and the communication protocol, including the hardware offload instruction set, can be triggered by this hardware engine. The second engine is a dedicated preprocessing hardware engine integrated into the RNIC, which can process DWQE (Device Work Queue Element) at hardware speed and free up the main processing unit resources of the CPU and RNIC. This reduces the number of data transfers during RDMA data transfer and reduces data latency. In addition, by directly offloading the RDMA data transfer task to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transfer task is decoupled from CPU / GPU resources. Therefore, it provides a high-efficiency and low-latency RDMA data transfer solution that does not rely on CPU / GPU resources for data transfer-intensive scenarios such as the MoE model.

[0013] It should be understood that the description in this section is not intended to identify key or essential features of the embodiments of this disclosure, nor is it intended to limit the scope of this disclosure. Other features of this disclosure will become readily apparent from the following description. Attached Figure Description

[0014] Other features, objects, and advantages of this disclosure will become more apparent from the following detailed description of non-limiting embodiments with reference to the accompanying drawings:

[0015] Figure 1 This is an exemplary system architecture to which this disclosure can be applied;

[0016] Figure 2 A block diagram of an RDMA data transmission network system architecture provided for embodiments of this disclosure;

[0017] Figure 3 A flowchart of an RDMA data transmission method provided for embodiments of this disclosure;

[0018] Figure 4 A block diagram of another RDMA data transmission network system architecture provided for embodiments of this disclosure;

[0019] Figure 5 A flowchart of another RDMA data transmission method provided for embodiments of this disclosure;

[0020] Figure 6 A block diagram of yet another RDMA data transmission network system architecture provided for embodiments of this disclosure;

[0021] Figure 7 This is a schematic diagram of the structure of an electronic device suitable for performing an RDMA data transmission method, provided for embodiments of the present disclosure. DETAILED DESCRIPTION

[0022] Exemplary embodiments of the present disclosure are described below with reference to the accompanying drawings, which include various details of implementations of the present disclosure to assist in understanding them. These should be considered as merely exemplary. Thus, those of ordinary skill in the art will recognize various changes and modifications of the embodiments described herein, which do not depart from the scope and spirit of the present disclosure. Also, descriptions of known functions and constructions are omitted for clarity and conciseness. Note that the embodiments in the present disclosure and the features in the embodiments can be combined with each other as long as there is no conflict.

[0023] In the technical solutions of the present disclosure, the collection, storage, use, processing, transmission, provision and disclosure of user personal information involved in the technical solutions comply with relevant laws and regulations and do not violate public order and good customs.

[0024] Figure 1 An exemplary system architecture 100 is shown, which can apply embodiments of the RDMA data transmission method, network device, system, electronic device and computer readable storage medium of the present disclosure.

[0025] As shown in Figure 1 The system architecture 100 can include a server 110, an RNIC 120 and an xPU 130, wherein the server 110 can include a CPU (Central Processing Unit), a host memory, etc., and can provide various services through various built-in applications. For example, the server 110 can provide a running environment and resource management for the xPU 130, allocate memory, and manage its communication with the RNIC (Remote Direct Memory Access Network Interface Controller) 120.

[0026] In the present disclosure, the "x" in xPU (x Processing Unit, which can be referred to as a general-purpose processor) is a wildcard, and the xPU can be regarded as a general term for various types of processors. For example, the xPU 130 can include at least one of a CPU, a GPU (Graphics Processing Unit), a TPU (Tensor Processing Unit), an NPU (Neural Processing Unit), a DPU (Data Processing Unit), a VPU (Vision Processing Unit), a QPU (Quantum Processing Unit), and an APU (Accelerated Processing Unit). In addition, the xPU 130 is generally represented as hardware, and in special scenarios (for example, simulation scenarios), the xPU 130 can also be represented as software or a software running product, and the present disclosure does not make specific limitations here.

[0027] In addition, the system architecture 100 can include multiple xPUs, such as a first xPU 1311, a second xPU 1312, a third xPU 1313, a fourth xPU 1314, and a fifth xPU 1315, etc. in the form of a graphics card. It should be noted that, Figure 1 The form of the RNIC 120 is simplified in the middle, and only a small number of xPUs 130 are shown as examples, but a person skilled in the art can set the number, type, and connection relationship of the server 110, the RNIC 120, and the xPU 130 in the system architecture 100 according to actual needs, and the present disclosure does not make limitations here.

[0028] The RNIC 120 can be used to provide communication between the server 110 and the xPU 130, and the RNIC 120 can include various connection types, and the present disclosure does not make limitations on the hardware form, interface type, and protocol support adopted by the RNIC 120.

[0029] In addition, a user can use a terminal device to interact with the server 110 through, for example, the RNIC 120 to receive or send messages, etc. The server 110 and the xPU 130 and the terminal device can all install various applications for realizing information communication between the two, such as training task issuing applications, training strategy optimization applications, instant messaging applications, etc.

[0030] The terminal device and the server 110 can be hardware or software. When the terminal device is hardware, it can be various electronic devices with a display screen, including but not limited to a smartphone, a tablet computer, a laptop computer, a desktop computer, and the like; when the terminal device is software, it can be installed in the above-mentioned electronic devices, and can be implemented as multiple software or software modules, or as a single software or software module, which is not specifically limited herein. When the server 110 is hardware, it can be implemented as a distributed server cluster composed of multiple servers, or as a single server; when the server 110 is software, it can be implemented as multiple software or software modules, or as a single software or software module, which is not specifically limited herein. The graphic processors constituting the xPU 130 are generally hardware, but can also be software or software running products in special scenarios (for example, simulation scenarios), which is not specifically limited herein.

[0031] The server 110 can provide various services through various built-in applications, for example, the server 110 can include at least one of a file server, a database server, a mail server, a Web server, an application server, a game server, IaaS (Infrastructure as a Service), PaaS (Platform as a Service), SaaS (Software as a Service), and an AI training server, wherein the file server can centrally store and share files (for example, an enterprise internal document library); the database server can run a database management system such as MySQL, PostgreSQL, etc.; the mail server can process the sending, receiving, and storage of mails; the Web server can host websites and Web applications (for example, Apache, Nginx); the application server can run middleware; the game server can process the real-time interaction logic of multiplayer online games; the IaaS can provide virtual machines, storage, network resources; the PaaS can provide development platforms and tools; the SaaS can deliver software over the network; and the AI training server can be equipped with GPU / TPU to accelerate deep learning model training, etc.

[0032] Figure 2 A block diagram of an RDMA data transmission network system architecture 100 provided for an embodiment of the present disclosure.

[0033] As shown in FIG. 1, the RDMA data transmission network system architecture 100 provided for an embodiment of the present disclosure includes a terminal device 100, a server 110, and an xPU 130. Figure 1 and Figure 2As shown, the system architecture 100 can include a network device, where the network device can include an RNIC 120 and an xPU 130. The xPU 130 can include a first engine 131, and the RNIC 120 can include a second engine 121 in communication with the first engine 131, a RDMA engine 122 in communication with the second engine 121, and a WQE Buffer 123.

[0034] Further, the xPU 130 can also include a core computing unit 132 and a memory management unit 133, where the core computing unit 132 and the memory management unit 133 of different types of xPU 130 can be different. For example, in the case that the xPU 130 includes a GPU, its core computing unit 132 can include a SM (Streaming Multiprocessor); in the case that the xPU 130 includes a CPU, its core computing unit 132 can include an ALU (Arithmetic Logic Unit), a CU (Control Unit), etc.

[0035] In addition, when the xPU 130 executes a program, the core computing unit 132 generates a virtual address, and the memory management unit 133 can be responsible for converting the virtual address into a physical address. Alternatively, in the case that the xPU 130 includes a GPU or a CPU, its memory management unit 133 can include an MMU (Memory Management Unit).

[0036] The RDMA engine 122 of the RNIC 120 can perform various data processing, where the data processing can include data transmission, memory access, and queue management, etc.

[0037] It should be understood that Figure 1 and Figure 2 the number of servers, networks, network devices, etc. in the above-mentioned figures is merely illustrative. Depending on the implementation needs, there can be any number of servers, networks, and network devices.

[0038] Figure 3 A flowchart of an RDMA data transmission method provided for an embodiment of the present disclosure, where reference is made to Figure 3 , the flow 200 can include the following steps:

[0039] Step 201: the first engine assembles a work element WE based on hardware offloading of an asynchronous copy instruction set, and transmits the WE to the second engine;

[0040] This step aims to enable the communication protocol, including the hardware unload instruction set, to be triggered by a hardware engine (the first engine). This hardware engine can be directly integrated within the xPU and can assemble its own Work Elements (WEs). Through methods such as PCIe BAR space mapping, the assembled WEs can be directly output to the second engine located in the RNIC. This reduces the dependence of RDMA data transfer on CPU or GPU performance. Furthermore, it reduces the number of data transfers during RDMA data transfer and decreases data latency during the aforementioned transfer processes.

[0041] Specifically, the hardware offloading asynchronous copy instruction set may include multi-level synchronization control data transmission instructions, which may include thread-level synchronization control data transmission instructions, thread group-level synchronization control data transmission instructions, memory block-level synchronization control data transmission instructions, and global-level synchronization control data transmission instructions.

[0042] For example, the hardware unloading asynchronous copy instruction set may include:

[0043] ibcp.async(dst.addr, src.addr, length);

[0044] ibcp.async.wait_thread;

[0045] ibcp.async.wait_wrap;

[0046] ibcp.async.wait_block;

[0047] ibcp.async.wait_all.

[0048] Among them, `ibcp.async(dst.addr, src.addr, length)` is the instruction used to initiate asynchronous data transmission, which is used to initiate asynchronous data transmission from the source address `src.addr` to the destination address `dst.addr` with a length of `length`; `ibcp.async.wait_thread` is a thread-level wait synchronization instruction; `ibcp.async.wait_wrap` is a thread group-level wait synchronization instruction; `ibcp.async.wait_block` is a memory block-level wait synchronization instruction; and `ibcp.async.wait_all` is a global wait synchronization instruction.

[0049] For example, ibcp.async(dst.addr, src.addr, length) initiates an asynchronous one-sided RDMA data transfer that writes data from local, e.g., xPU memory (src.addr) to remote device memory (dst.addr) without waiting for the transfer to complete. This instruction can be executed by a dedicated hardware engine, avoiding the use of xPU resources.

[0050] For example, ibcp.async.wait_thread waits for all RDMA data transfers initiated by ibcp.async(dst.addr, src.addr, length) in the current xPU thread to complete, with synchronization granularity at the thread level. After initiating multiple RDMA data transfers in the same thread, it is ensured that these transfers are completed before proceeding to subsequent operations.

[0051] For example, ibcp.async.wait_wrap waits for all RDMA data transfers initiated by ibcp.async(dst.addr, src.addr, length) in the current xPU thread group to complete, with synchronization granularity at the thread group level. After initiating multiple RDMA data transfers in the same thread group, it is ensured that these transfers are completed before proceeding to subsequent operations.

[0052] For example, ibcp.async.wait_block waits for all RDMA data transfers initiated by ibcp.async(dst.addr, src.addr, length) in the current xPU storage block to complete, with synchronization granularity at the storage block level. After initiating multiple RDMA data transfers in the same storage block, it is ensured that these transfers are completed before proceeding to subsequent operations.

[0053] For example, ibcp.async.wait_all waits for all RDMA data transfers initiated by ibcp.async(dst.addr, src.addr, length) in the xPU to complete, with synchronization granularity at the global level. After initiating all RDMA data transfers, it is ensured that all transfers are completed before proceeding to subsequent operations.

[0054] Thus, in some embodiments of the present disclosure, by including the multi-level synchronization control data transmission instruction in the hardware offloading asynchronous copy instruction set, the RDMA data transmission task is directly offloaded to the dedicated hardware engine, and the computing resources of the xPU are completely released. After initiating the RDMA data transmission, the xPU can continue to perform the computing task and can be synchronized on demand by the multi-level synchronization control data transmission instruction, maximizing the utilization of computing resources. In addition, the multi-level synchronization of threads, thread groups, memory blocks, and the like can adapt to different scene requirements.

[0055] Optionally, in some embodiments of the present disclosure, the first engine assembles the work element WE based on the hardware offloading asynchronous copy instruction set can include: receiving an RDMA operation request; assembling the RDMA operation request into a WE based on the hardware offloading asynchronous copy instruction set; and writing the WE into a second register, wherein the second register is located in the RNIC and can communicate with the second engine.

[0056] Specifically, the first engine can receive an RDMA operation request, such as an RDMA Copy Async message, issued by any xPU in the system architecture through a NOC (Network on Chip). The RDMA operation request is assembled into a WE based on the hardware offloading asynchronous copy instruction set. The NOC is a communication architecture inside an integrated circuit, used to connect multiple components (such as processors, memories, etc.) within a chip, and can achieve efficient data transmission and task scheduling.

[0057] The WE can include at least one of remote information Remote info, local information Local info, an operation code Opcode, a QPN (Queue Pair Number), and a tag TAG, wherein the TAG is used to mark the thread identifier ID or the block ID of the source of the RDMA operation request.

[0058] For example, the Remote info can be represented as (Remote VA, Rkey); the Local info can be represented as (localVA, lkey, length), wherein the Remote VA is the virtual address of the target memory on the remote device, the Rkey is the remote memory access key; the local VA is the virtual address of the memory on the local device; the lkey is the local memory access key; and the length is the length of the data transmission in bytes. The WE information collectively constitutes the basic parameters of the RDMA data transmission operation, which can ensure the efficient and safe execution of the RDMA data transmission.

[0059] Optionally, the way of writing the WE into the second register can include a device direct PCIe P2P way. The PCIe P2P way (Peer to Peer) refers to direct communication between two devices or components without the need for a processor or memory to relay. This transmission method significantly improves the processing speed of RDMA data transmission tasks and the overall performance of the system. For example, the PCIe P2P way is a hardware-level direct communication that bypasses the processor and memory; it achieves asynchronous operation optimization, where the hardware engine (second engine) communicating with the second register can automatically process tasks (e.g., storing WE in the WQE Buffer) after writing WE into the second register without waiting for software instructions; in addition, the direct hardware path can reduce data transmission delay.

[0060] Step 202: The second engine stores the WE in the WQE Buffer and transmits it to the RDMA engine;

[0061] On the basis of step 201, this step aims to use a hardware engine (second engine) to process DWQE at hardware speed and release CPU and RNIC main processing unit resources. On the basis of reducing single operation delay and improving network device throughput, the RDMA data transmission task is directly offloaded to a dedicated hardware engine through a hardware asynchronous instruction set, which realizes the decoupling of the RDMA data transmission task and the CPU / GPU resources, and thus provides a high-efficiency and low-delay RDMA data transmission solution for data transmission intensive scenarios such as MoE models without relying on CPU / GPU resources.

[0062] Specifically, the RDMA protocol is a network protocol that allows computers to directly access remote computer memory, commonly used in high-performance computing (HPC), data centers, storage networks, and other scenarios. RDMA can achieve high-speed data transmission between hosts without the intervention of the operating system kernel, thus having extremely low latency and high bandwidth. When using the RDMA protocol, the sender (i.e., the communication initiator) directly writes data into the memory of the receiver (usually referred to as the communication receiver) without the intervention of the operating system. This direct memory access reduces the intermediate layers in the data transmission process, thereby improving data transmission efficiency.

[0063] However, taking data transmission-intensive scenarios such as the MoE model as an example, each sample in the model can activate multiple experts, which leads to a large demand for transmitting numerous scattered small data blocks. In the RDMA data transmission process, each task needs to generate a WQE (Work Queue Element), trigger data processing through a DB (Doorbell), and actively check the completion events in the CQ (Completion Queue) through a thread loop.

[0064] In some embodiments, CPU threads can be used to encapsulate the task units to be sent (e.g., network requests) into WQEs and submit them to the send queue for processing. However, in RDMA data transmission, the number of QPs is relatively large, or in other words, the number of basic units used to manage send / receive queues in technologies such as RDMA is large. To meet the needs of transmitting a large number of dispersed small data blocks, and to implement the tasks of generating WQEs, triggering DBs, and checking CQs, a large amount of CPU resources will be consumed. Therefore, with the overhead of data transmission increasing exponentially, system performance becomes a bottleneck for scaling up the model.

[0065] In other embodiments, the multi-threaded parallel capabilities of the GPU can be used to complete the processing of multiple QP tasks. For example, the GPU can use multi-threading to complete tasks such as generating WQE, triggering DB, and checking CQ. This decouples the RDMA data transfer task from CPU performance; however, it can consume a large amount of GPU SM computing resources when the data transfer overhead increases exponentially.

[0066] Therefore, in the embodiments of this disclosure, the first engine is directly integrated inside the xPU, and the communication protocol, including the hardware offload instruction set, can be triggered by this hardware engine; the second engine is a dedicated preprocessing hardware engine, which is integrated in the RNIC and can process DWQE at hardware speed, freeing up the main processing unit resources of the CPU and RNIC. This reduces the number of data transfers during RDMA data transfer and reduces data latency; in addition, by directly offloading the RDMA data transfer task to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transfer task is decoupled from CPU / GPU resources, thus providing a high-efficiency and low-latency RDMA data transfer solution that does not rely on CPU / GPU resources for data transfer-intensive scenarios such as the MoE model.

[0067] In particular, in some embodiments of the present disclosure, the second engine storing the WE in the WQE Buffer and sending to the RDMA engine can include: parsing the WE and storing the WE in the WQE Buffer; and transmitting a plurality of WE stored in the WQE Buffer to the host memory of the network device based on the occupancy of the WQE Buffer being greater than or equal to a preset threshold.

[0068] The second engine can receive the WE from the PCIe bus, for example, and can verify the integrity of the WE information after receiving the WE. The received WE information is checked for completeness, data loss or damage. For example, whether the length in the WE information is as expected, whether the checksum in the WE information is correct, and the like. If the WE information is incomplete or has errors, the WE can be discarded and the error information can be recorded.

[0069] In addition, the second engine can also parse the WE and parse each field in the WE to determine the specific requirements of the operation. For example, parsing Remote info (Remote VA, Rkey), local Info (local VA, lkey, length), Opcode, QPN, TAG and the like.

[0070] After parsing the WE, the second engine can extract key information in the WE. The key information is extracted from the parsed fields to provide a basis for subsequent processing and decision-making. For example, the operation type is determined according to Opcode. Optionally, the operation type can include reading, writing, atomic operation, and the like; and the queue to which the operation belongs is determined according to QPN.

[0071] According to the information obtained by parsing, the second engine can also allocate cache space in the WQE Buffer for the WE. Based on the occupancy of the WQE Buffer being greater than or equal to a preset threshold, a plurality of WE stored in the WQE Buffer is transmitted to the host memory of the network device. In other words, if the on-chip WQE Buffer is full, the WQE can be pushed into the host memory (such as the WQE external buffer 111 of the host memory shown in FIG. 1) of the network device. This hierarchical storage can guarantee performance through on-chip cache (for example, WQE Buffer), provide capacity redundancy through external memory (for example, WQE external buffer 111), and ultimately realize an efficient and reliable WQE processing pipeline. Figure 2

[0072] In addition, after parsing the WE, the second engine can schedule the processing order. For example, the processing order of the WE is determined according to the priority of the operation, the queue state and the like. High-priority WE is processed first. This hardware arbitration mechanism automatically assigns task priorities, which can avoid queue blocking.

[0073] ​Optionally, the second engine storing the WE in the WQE Buffer and sending the WE to the RDMA engine can further include generating a DB, wherein the DB can include at least one of the queue pair information to be processed, the start address of the plurality of WE, and the number of the plurality of WE, and transmitting the DB to the RDMA engine, which can trigger the RDMA engine to perform the data processing, wherein the RDMA engine can read the WE from the WQE Buffer or the WQE external buffer in the host memory.

[0074] The generation and transmission of the DB itself is a sign of the start of the task. When the DB is written into a specific memory area (such as an on-chip cache or a designated location in the host memory), the RDMA engine automatically starts the processing flow after detecting the presence of the DB through hardware logic, without the intervention of software or additional interrupt signals. In addition, the DB can be located in a shared area of the on-chip cache or the host memory, which can be directly accessed by the RDM engine, thus avoiding data copying overhead. After completing the current data processing, the RDM engine can immediately obtain the next data processing through the doorbell mechanism of the DB, without waiting for an interrupt response.

[0075] Step 203: The RDM engine performs data processing based on the WE, which can include data transmission, memory access, and queue management.

[0076] This step aims to describe that the RDMA engine achieves low-latency, high-bandwidth, and low-xPU resource occupation in RDMA data transmission processing by performing data processing tasks including data transmission, memory access, and queue management. The core advantage is to completely offload complex network operations (such as transmission control and memory management) to hardware, so that the CPU can focus on business logic, thereby significantly improving the performance of distributed systems.

[0077] Optionally, the RDMA engine can achieve high-speed data transmission between hosts through direct memory access technology without the intervention of the operating system kernel, thus having extremely low latency and high bandwidth. For example, the WE contains a source address and a target address, and the RDMA engine can directly move data from the source address to the send buffer of the RNIC through direct memory access technology without copying.

[0078] In addition, the RDMA engine can support multiple types of transmission, such as Send / Receive, Read / Write, etc., where Send / Receive can be understood as bidirectional data transfer between local or remote nodes, and Read / Write can be understood as direct access to the memory of a remote node. This can eliminate the context switching and interrupt overhead of xPU participation, and reduce the data transmission delay to the microsecond level.

[0079] The RDMA engine supports direct memory access across nodes by mapping virtual addresses of remote nodes to physical addresses through a Memory Window technology.

[0080] In addition, each queue pair (QP) can include an independent send queue (SQ) and receive queue (RQ), so that the RDMA engine supports concurrent processing of multiple task flows.

[0081] Step 204: The first engine receives feedback of data processing transmitted by the second engine.

[0082] This step aims to feed back the data processing completion message based on the RDMA operation request to the xPU that issues the RDMA operation request. Alternatively, the first engine can transmit the feedback of the data processing to the xPU that issues the RDMA operation request through the NOC. The NOC is a communication architecture inside an integrated circuit, which is used to connect multiple components (such as processors, memories, etc.) within a chip, and can achieve efficient data transmission and task scheduling.

[0083] Specifically, after completing the above data processing, the RDMA engine can generate a completion event (CE) message; the second engine receives the CE message and writes the received CE message into the first register, wherein the first register is located in the xPU and can communicate with the first engine. This can ensure the reliability of RDMA data transmission.

[0084] The first register and the second register described above can be understood as part of a BAR (Base Address Register) register. The first engine, the second engine, and the RDMA engine in the system architecture can be located in different hardware modules or processing units (for example, xPU), and can communicate with each other through the NOC or shared registers (for example, BAR registers). This avoids the first engine, the second engine, and the RDMA engine directly transferring, which can require complex hardware connections, reduces the complexity of design, and can ensure that different hardware modules or processing units work independently, reducing mutual interference.

[0085] In addition, the BAR register is a register reported by a PCIe device (for example, RNIC, xPU, etc.) to a server in an enumeration phase, which is used to declare the memory space (for example, configuration space, I / O space, memory mapping space, etc.) required by the device. The server can allocate a physical address for the BAR register, so that as part of the memory-mapped I / O, the BAR register provides a standardized interface, which can complete data interaction through simple read and write operations, reducing design difficulty and power consumption.

[0086] Optionally, the received CE message can be written to the first register via PCIe P2P. This transmission method significantly improves the processing speed of RDMA data transfer tasks and the overall system performance. For example, PCIe P2P is a direct hardware-level communication method that bypasses the processor and memory; it achieves asynchronous operation optimization. After the CE is written to the first register, the hardware engine (first engine) communicating with the first register can automatically process the task (e.g., parse the CE) without waiting for software instructions; in addition, the direct hardware path can reduce data transfer latency.

[0087] Alternatively, the CE message may include at least one of the following: opcode, queue pair number QPN, and tag, where the TAG is used to identify the thread identifier ID of the RDMA operation request or the block ID of the source. Together, the CE information constitutes the basic parameters of the RDMA data transfer operation, ensuring efficient and secure execution of RDMA data transfer.

[0088] After receiving the CE message, the first engine can generate a task completion message based on the CE message and pass it to the NOC, which is then transmitted to the xPU that issued the RDMA operation request.

[0089] Therefore, in the RDMA data transfer scheme provided by the embodiments of this disclosure, the first engine is directly integrated inside the xPU, and the communication protocol including the hardware offload instruction set can be triggered by this hardware engine; the second engine is a dedicated preprocessing hardware engine, which is integrated in the RNIC, and can process DWQE at hardware speed, thereby freeing up the main processing unit resources of the CPU and RNIC. This reduces the number of data transfers during RDMA data transfer and reduces data latency; in addition, by directly offloading the RDMA data transfer task to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transfer task is decoupled from CPU / GPU resources, thus providing a high-efficiency and low-latency RDMA data transfer solution that does not rely on CPU / GPU resources for data transfer-intensive scenarios such as the MoE model.

[0090] Figure 4 A block diagram of another RDMA data transmission network system architecture 100 provided for embodiments of this disclosure. Figure 5 A flowchart of another RDMA data transmission method provided for embodiments of this disclosure.

[0091] like Figure 2 and Figure 4 As shown, the system architecture 100 may include multiple network devices, such as a first network device and a second network device, wherein the first network device may include server A and xPUA, etc., and the second network device may include server B and xPUB, etc.

[0092] In addition, it should be noted that the system architecture 100 can further include the rest of the network devices in addition to the first network device and the second network device, Figure 4 The network devices in the system architecture 100 are shown as an example in a small number, and a person skilled in the art can set the number, type and connection relationship of the network devices in the system architecture 100 according to actual needs, and the present disclosure does not limit this.

[0093] Each of the plurality of network devices can include an RNIC and an xPU. The xPU can include a first engine, and the RNIC can include a second engine in communication with the first engine, an RDMA engine in communication with the second engine, and a WQE Buffer.

[0094] In addition, in each of the plurality of network devices, the xPU can further include a core computing unit and a memory management unit, wherein the types of the core computing unit and the memory management unit of different types of xPU can be different. For example, in the case where the xPU includes a GPU, the core computing unit thereof can include an SM; in the case where the xPU includes a CPU, the core computing unit thereof can include an arithmetic logic unit, a control unit, etc.

[0095] In addition, when the xPU executes a program, the core computing unit generates a virtual address, and the memory management unit can be responsible for converting the virtual address into a physical address. Alternatively, in the case where the xPU includes a GPU or a CPU, the memory management unit thereof can include an MMU.

[0096] In each of the plurality of network devices, the RDMA engine of the RNIC can perform a plurality of data processing, wherein the data processing can include data transmission, memory access, and queue management, etc.

[0097] It should be understood that Figure 4 The number of servers, networks, network devices, etc. in the system architecture 100 is only illustrative. According to the needs of implementation, there can be any number of servers, networks and network devices.

[0098] Hereinafter, the RDMA data transmission method will be described by taking the first network device accessing the memory of the second network device as an example. It should be noted that the RDMA data transmission method described in the foregoing can be completely or partially applicable to the RDMA data transmission method described hereinafter, and the related or similar contents thereof will not be described again. However, a person skilled in the art can understand that the structural features, method processes, implementation principles and technical effects achieved by the RDMA data transmission methods described in the two are similar.

[0099] Referring to Figure 5 , the flow 300 can include the following steps:

[0100] Step 301: The first engine of the first network device assembles the working element WE based on the hardware offload asynchronous copy instruction set and transmits the WE to the second engine;

[0101] This step aims to enable the communication protocol, including the hardware unload instruction set, to be triggered by a hardware engine (the first engine). This hardware engine can be directly integrated within the xPU and can assemble its own Work Elements (WEs). Through methods such as PCIe BAR space mapping, the assembled WEs can be directly output to the second engine located in the RNIC. This reduces the dependence of RDMA data transfer on CPU or GPU performance. Furthermore, it reduces the number of data transfers during RDMA data transfer and decreases data latency during the aforementioned transfer processes.

[0102] Furthermore, in some embodiments of this disclosure, before performing step 301, the RDMA data transmission method may further include: creating an interface corresponding to at least a portion of the RDMA's MR (Memory Region), the interface information including a unique identifier UniqueID; based on the interface, allocating the MR to multiple network devices and establishing MR page tables for the network devices, wherein the memory access credentials mKey for the multiple MR page tables are all UniqueID; the multiple network devices implement memory handle exchange; binding the exchanged memory handle to the information of the WE and establishing a virtual address mapping relationship to facilitate remote access through the memory handle and UniqueID.

[0103] Specifically, with RNIC hardware support, a new RDMA memory region can be provided to create the aforementioned interface. For example, the interface can be created using the instruction `rdma_register_mr_unique(va, length, uniqueID)`. `UniqueID` is an ID used to identify the aggregate communication; it is synchronized across all nodes in the cluster during communication initialization. The nodes described here can be understood as network devices in the system architecture.

[0104] A memory region can be understood as an abstraction of a contiguous block of memory in RDMA. The interface allows users to create a MR by specifying a virtual address (va), length, and unique identifier (uniqueID). The uniqueID ensures that in cluster communication, each node can identify and synchronize specific communication tasks.

[0105] Taking the first network device as an example, the xPUA can apply for an HBM (High Bandwidth Memory) through a cudaMalloc function, create a memory region corresponding to the first network device by using the above interface, establish an MR page table of the first network device, and make the access credentials mKey of the MR page table all be UniqueIDs. After the memory region of the first network device is created and the MR page table is established, the operation of RDMA data transmission can be located to the memory region of the first network device, and the mKey can be used as a key credential for the memory access.

[0106] Similarly, the memory region and the MR page table can also be created in the second network device and the remaining network devices. In this way, the operation of RDMA data transmission can also be located to the memory region of the second network device and the memory region of the remaining network devices, and the mKey can be used as a key credential for the memory access. In addition, the multiple network devices in the system architecture can also access each other, and the same UniqueID can ensure the communication association between the multiple network devices.

[0107] Optionally, the application program of the second network device can assemble a memory handle (xPUB Memory Handle) and interact with the remaining network devices (for example, the first network device) through a Socket or other ways, and pass the memory handle to the remaining network devices. The memory handle can include key information for accessing remote memory, such as a virtual address (xPUB va) and a length (length) of the second network device.

[0108] Similarly, the application program of the first network device can assemble a memory handle (xPUA Memory Handle) and interact with the remaining network devices (for example, the second network device) through a Socket or other ways, and pass the memory handle to the remaining network devices. The memory handle can include key information for accessing remote memory, such as a virtual address (xPUA va) and a length (length) of the first network device.

[0109] The network device can bind the obtained memory handle after the exchange to the information of the WE, and establish a virtual address mapping relationship, so as to initiate remote access through the memory handle and the UniqueID.

[0110] For example, the network device can bind the obtained memory handle after the exchange to a QP, and specify which QP to complete the access to the network device corresponding to the memory handle. In addition, the QP is edited in the information of the WE. In this way, the WE can specify the communication path of the network device access, and ensure that the data can be accurately transmitted between the network devices.

[0111] Optionally, taking the first network device as an example, the first network device can apply for a local reserved virtual address (for example, Local Reserved VA) space locally through a CudaMallocReserved function, establish mapping information for it, and make it point to a memory handle (for example, an xPUB Memory Handle of the second network device) of the remaining network device. The mapping information can be managed by the MMU of the first network device, or can be managed in the form of a table item. In other words, the network device can reserve a virtual address space locally as a local reserved virtual address. After the mapping is established, when the network device accesses the local reserved virtual address, the local reserved virtual address can be converted into the actual address of the network device to be accessed and related communication information through the mapping information, so as to realize remote memory access.

[0112] On this basis, taking the first network device as an example, the xPU of the first network device can start a kernel program (Kernel Launch), the kernel program is executed by multiple threads in parallel, and a hardware offload asynchronous copy instruction set is used to initiate remote memory access. The UniqueID is injected into the process context, and can be transmitted to the MMU and RNIC through the hardware offload asynchronous copy instruction set in the subsequent process, and is used to associate the metadata of the remote memory access.

[0113] Taking ibcp.async(xPUB Reserved VA, Local VA, length) as an example, the hardware offload asynchronous copy instruction set is transmitted to the MMU, and the MMU can complete address translation, and convert the reserved virtual address xPUB Reserved VA of the second network device into the corresponding virtual address xPUB VA and QPN information of the second network device. Then, the MMU transmits the virtual address xPUA VA, xPUB VA, QPN, length and UniqueID information of the first network device to the first engine.

[0114] Optionally, the hardware offload asynchronous copy instruction set can include ibcp.async(dst.addr, src.addr, length). When Dst.addr is xPUB Reseverd VA, it means that the memory pointed to by the xPUA local va of the local first network device is written (Write) to the memory of the remote second network device. When Src.addr is xPUB Reseverd VA, it means that the memory of the remote second network device xPUB Memory is read (Read) to the memory pointed to by the xPUA local va of the local first network device.

[0115] The UniqueID is used to identify the ID of the collective communication, and synchronization of all nodes in the cluster has been completed at the initialization of the communication. In other words, at the starting point of the RDMA data operation, by starting the kernel program and carrying the UniqueID, it can be ensured that the operation can be accurately associated with a specific communication task, avoiding confusion between different communication tasks.

[0116] The MMU is used to convert the virtual address into the physical address or the corresponding remote address information. Through address translation, the actual memory address of the network device can be determined, and the relevant operation information is transmitted to the first engine and the second engine, so as to prepare for subsequent RDMA data transmission.

[0117] The first engine assembles a work element WE based on the above information and a hardware offloaded asynchronous copy instruction set, and transmits the WE to the second engine.

[0118] Step 302: The second engine of the first network device stores the WE in the WQE Buffer and transmits it to the RDMA engine of the first network device;

[0119] On the basis of step 301, this step aims to use the hardware engine (the second engine) to process the DWQE at a hardware speed, and release the CPU and the main processing unit resource of the RNIC. On the basis of reducing the single operation delay and improving the throughput of the network device, the RDMA data transmission task is directly offloaded to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transmission task is decoupled from the CPU / GPU resource, and thus a high-efficiency and low-delay RDMA data transmission solution for a data transmission intensive scene such as the MoE model is provided without relying on the CPU / GPU resource.

[0120] Step 303: The RDMA engine of the first network device initiates a request to the second network device based on the WE;

[0121] This step aims to describe that the RDMA engine realizes the low-delay, high-bandwidth, and low-xPU resource occupied RDMA data transmission processing by performing the data processing task including data transmission, memory access, and queue management. The core advantage is that the complex network operation (for example, transmission control and memory management) is completely offloaded to the hardware, and the CPU focuses on the business logic, so that the performance of the distributed system is significantly improved.

[0122] Optionally, in some embodiments of this disclosure, the RDMA engine of the first network device may obtain the QPC (Queue Pair Context) corresponding to the queue pair identifier QPN of the WE based on the WE, and determine the operation type according to the QPC; the RDMA engine of the first network device may also obtain the memory handle and UniqueID of the second network device based on the WE; the RDMA of the first network device completes the packet encapsulation based on the WE and sends a request to the second network device.

[0123] Step 304: The second network device completes the above request and sends the feedback to the first engine of the first network device.

[0124] This step aims to send a data processing completion message based on the RDMA operation request back to the xPU that issued the RDMA operation request. Optionally, the first engine can transmit this data processing feedback to the xPU that issued the RDMA operation request via the NOC. The NOC is the communication architecture within an integrated circuit, used to connect multiple components within the chip (such as processors, memory, etc.), enabling efficient data transmission and task scheduling.

[0125] Therefore, in the RDMA data transfer scheme provided by the embodiments of this disclosure, the first engine is directly integrated inside the xPU, and the communication protocol including the hardware offload instruction set can be triggered by this hardware engine; the second engine is a dedicated preprocessing hardware engine, which is integrated in the RNIC, and can process DWQE at hardware speed, thereby freeing up the main processing unit resources of the CPU and RNIC. This reduces the number of data transfers during RDMA data transfer and reduces data latency; in addition, by directly offloading the RDMA data transfer task to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transfer task is decoupled from CPU / GPU resources, thus providing a high-efficiency and low-latency RDMA data transfer solution that does not rely on CPU / GPU resources for data transfer-intensive scenarios such as the MoE model.

[0126] Figure 6 A block diagram of another RDMA data transmission network system architecture 100 provided for embodiments of this disclosure.

[0127] Combination Figure 5 and Figure 6The following section will use a write operation as an example to describe the RDMA data transfer method. Specifically, the first network device can initiate a write operation request such as ibcp.async to write data from the local memory of the first network device to the remote memory of the second network device. It should be noted that since the content of the RDMA data transfer method described above is wholly or partially applicable to the RDMA data transfer method described below, related or similar content will not be repeated. However, those skilled in the art will understand that the structural characteristics, process, implementation principle, and technical effects of the RDMA data transfer method described herein are similar.

[0128] refer to Figure 5 Process 300 may include the following steps:

[0129] Step 301: The first engine of the first network device assembles the working element WE based on the hardware offload asynchronous copy instruction set and transmits the WE to the second engine;

[0130] This step aims to enable the communication protocol, including the hardware unload instruction set, to be triggered by a hardware engine (the first engine). This hardware engine can be directly integrated within the xPU and can assemble its own Work Elements (WEs). Through methods such as PCIe BAR space mapping, the assembled WEs can be directly output to the second engine located in the RNIC. This reduces the dependence of RDMA data transfer on CPU or GPU performance. Furthermore, it reduces the number of data transfers during RDMA data transfer and decreases data latency during the aforementioned transfer processes.

[0131] Furthermore, in some embodiments of this disclosure, before performing step 301, the RDMA data transmission method may further include: creating an interface corresponding to at least a portion of the RDMA memory region MR, the interface information including a unique identifier QPC; based on the interface, allocating the MR to multiple network devices and establishing MR page tables for the network devices, wherein the memory access credentials mKey for the multiple MR page tables are all UniqueID; the multiple network devices implement memory handle exchange; binding the exchanged memory handles to the information of the WE and establishing a virtual address mapping relationship to facilitate remote access through the memory handles and UniqueID.

[0132] In the process of performing step 301, the xPU of the first network device can start a kernel program KernelLaunch, which is executed by multiple threads in parallel and initiates remote memory access through hardware offloaded asynchronous copy instruction sets. The UniqueID is injected into the process context and can be subsequently transmitted to the MMU and RNIC through the hardware offloaded asynchronous copy instruction sets for associating the metadata of the remote memory access.

[0133] The UniqueID is used to identify the ID of the collective communication, and synchronization of all nodes in the cluster has been completed at the time of communication initialization. In other words, at the starting point of the RDMA data operation, by starting the kernel program and carrying the UniqueID, it can be ensured that the operation can be accurately associated to a specific communication task, avoiding confusion between different communication tasks.

[0134] The role of the MMU is to convert the virtual address into the physical address or the corresponding remote address information. Through address translation, the actual memory address of the network device can be determined, and the relevant operation information is transmitted to the first engine and the second engine for subsequent preparation work for RDMA data transmission.

[0135] Taking ibcp.async(xPUB Reserved VA, Local VA, length) as an example, the hardware offloaded asynchronous copy instruction set is transmitted to the MMU, and the MMU can complete address translation to convert the reserved virtual address xPUB Reserved VA of the second network device into the corresponding virtual address xPUB VA and QPN information of the second network device. Then, the MMU transmits the virtual address xPUA VA, xPUB VA, QPN, length and UniqueID information of the first network device to the first engine.

[0136] The first engine assembles a work element WE based on the above information and the hardware offloaded asynchronous copy instruction set, and transmits the WE to the second engine.

[0137] Step 302: The second engine of the first network device stores the WE in the WQE Buffer and transmits it to the RDMA engine of the first network device;

[0138] On the basis of step 301, the present step aims to process the DWQE at a hardware speed by using a hardware engine (second engine) and release the CPU and main processing unit resources of the RNIC. On the basis of reducing the single operation delay and improving the throughput of the network device, the RDMA data transmission task is directly unloaded to the dedicated hardware engine through the hardware asynchronous instruction set, the decoupling of the RDMA data transmission task and the resources of the CPU / GPU is realized, and thus a high-efficiency and low-delay RDMA data transmission solution for data transmission intensive scenarios such as the MoE model and the like is provided without relying on the CPU / GPU resources.

[0139] Step 303: The RDMA engine of the first network device initiates a request to the second network device based on the WE.

[0140] The present step aims to describe that the RDMA engine realizes the RDMA data transmission processing with low delay, high bandwidth and low xPU resource occupation by performing data processing tasks including data transmission, memory access and queue management. The core advantage lies in that the complex network operation (for example, transmission control and memory management) is completely unloaded to the hardware, and the CPU focuses on the business logic, so that the performance of the distributed system is significantly improved.

[0141] Specifically, the RDMA engine of the first network device can obtain the corresponding QPC based on the QPN, initiate the RDMA Write operation, and complete the encapsulation of the RDMA Write Request packet according to the hardware unloading asynchronous copy instruction set and the WE, and send it to the remote network device (for example, the second network device). For example, the data in the high-bandwidth memory of the second network device can be obtained according to the xPUB va, UniqueID (as mKey) and length information.

[0142] Step 304: The second network device completes the above request and feeds back to the first engine of the first network device.

[0143] The present step aims to feed back the data processing completion message based on the RDMA operation request to the xPU that initiates the RDMA operation request. Alternatively, the first engine can transmit the feedback of the data processing to the xPU that initiates the RDMA operation request through the NOC. The NOC is a communication architecture inside the integrated circuit, which is used to connect multiple components (such as processors, memories, etc.) in the chip, and can realize efficient data transmission and task scheduling.

[0144] For example, upon receiving the above-mentioned request packet (e.g., data write packet), the RDMA engine of the second network device can write the payload data (Payload) in the request packet into the memory of the second network device according to the remote endpoint transport header, UniqueID, and length information, where the remote endpoint transport header can include the virtual address (xPUB VA) of the second network device. By offloading the asynchronous copy instruction set and WE information through hardware, the second network device can be accurately located, and the xPU of the second network device can be bypassed, reducing the intervention and overhead of the xPU, thereby significantly improving the efficiency of RDMA data transmission and avoiding the xPU from becoming a bottleneck of data transmission.

[0145] The RDMA engine of the second network device can also return an ACK (Acknowledgment) to the RDMA engine of the first network device, thereby informing the first network device that the RDMA write request packet sent by the first network device has been successfully received and processed. This can ensure the reliability of RDMA data transmission.

[0146] Upon receiving the ACK, the RDMA engine of the first network device can generate a CE message according to the QP and WQE, and deliver the CE message to the second engine. The second engine can assemble the received CE message, for example, to make the format and content of the message more standardized and unified. Then, the second engine can deliver the CE message to the first engine in a P2P manner. By generating and delivering the CE message, the RDMA engine of the first network device can timely inform the hardware engine that the operation has been completed. This enables the hardware engine to timely perform subsequent processing, such as updating state information, releasing resources, or triggering new tasks. In addition, the P2P manner avoids the delay and bandwidth competition problems caused by message transmission through intermediate nodes or shared buses. Through the P2P manner, the CE message can be quickly delivered from the second engine to the first engine, improving the response speed and overall performance of the network device.

[0147] Therefore, in the RDMA data transfer scheme provided by the embodiments of this disclosure, the first engine is directly integrated inside the xPU, and the communication protocol including the hardware offload instruction set can be triggered by this hardware engine; the second engine is a dedicated preprocessing hardware engine, which is integrated in the RNIC, and can process DWQE at hardware speed, thereby freeing up the main processing unit resources of the CPU and RNIC. This reduces the number of data transfers during RDMA data transfer and reduces data latency; in addition, by directly offloading the RDMA data transfer task to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transfer task is decoupled from CPU / GPU resources, thus providing a high-efficiency and low-latency RDMA data transfer solution that does not rely on CPU / GPU resources for data transfer-intensive scenarios such as the MoE model.

[0148] Refer again Figure 2 As an implementation of the aforementioned RDMA data transmission method, embodiments of this disclosure provide an RDMA data transmission network device, which can be connected to... Figure 3 Corresponding to the method shown, this network device can be specifically applied to various electronic devices.

[0149] like Figure 2 As shown, the RDMA data transmission network device provided in the embodiments of this disclosure may include an RNIC 120 and an xPU 130. The xPU 130 may include a first engine 131, and the RNIC 120 may include a second engine 121 communicating with the first engine 131, an RDMA engine 122 communicating with the second engine 121, and a WQE buffer 123. The first engine 131 is configured to assemble working elements (WEs) based on a hardware offload asynchronous copy instruction set and transmit the WEs to the second engine; the second engine 121 is configured to store the WEs in the WQE buffer 123 and transmit them to the RDMA engine 122; the RDMA engine 122 is configured to perform data processing based on the WEs and transmit feedback of the data processing to the first engine 131 via the second engine 121. The data processing includes data transmission, memory access, and queue management.

[0150] Further, the xPU 130 can also include a core computing unit 132 and a memory management unit 133, where the types of the core computing unit 132 and the memory management unit 133 of different types of xPU 130 can be different. For example, in the case where the xPU 130 includes a GPU, the core computing unit 132 thereof can include a streaming multiprocessor (SM); in the case where the xPU 130 includes a CPU, the core computing unit 132 thereof can include an arithmetic logic unit (ALU), a control unit (CU), etc.

[0151] In addition, when the xPU 130 executes a program, the core computing unit 132 generates a virtual address, and the memory management unit 133 can be responsible for converting the virtual address into a physical address. Alternatively, in the case where the xPU 130 includes a GPU or a CPU, the memory management unit 133 thereof can include a memory management unit (MMU).

[0152] Alternatively, the xPU 130 can include at least one of a CPU, a GPU, a TPU, a NPU, a DPU, a VPU, a QPU, and an APU. In addition, the xPU 130 can generally be hardware, and in special scenarios (for example, an emulation scenario), can also be software or a software running product, which is not specifically limited in the present disclosure.

[0153] Alternatively, the RDMA data transmission network device can include a plurality of xPU 130. In other words, a person skilled in the art can set the number, type, and connection relationship of the RNIC 120 and the xPU 130 in the RDMA data transmission network device according to actual needs, which is not limited in the present disclosure.

[0154] In some embodiments of the present disclosure, the hardware offloading asynchronous copy instruction set can include multi-level synchronization control data transmission instructions, where the multi-level synchronization control data transmission instructions can include thread-level synchronization control data transmission instructions, thread group-level synchronization control data transmission instructions, memory block-level synchronization control data transmission instructions, and global-level synchronization control data transmission instructions. By including the multi-level synchronization control data transmission instructions in the hardware offloading asynchronous copy instruction set, the RDMA data transmission task is directly offloaded to a dedicated hardware engine, and the computing resources of the xPU are completely released. After initiating the RDMA data transmission, the xPU can continue to execute the computing task and can be synchronized on demand through the multi-level synchronization control data transmission instructions, maximizing the utilization of computing resources. In addition, multi-level synchronization such as thread, thread group, memory block, and global is provided to adapt to different scenario requirements.

[0155] Optionally, the first engine 131 can be configured to assemble the work element WE based on the hardware offloaded asynchronous copy instruction set. For example, the first engine 131 can receive a RDMA operation request; assemble the RDMA operation request into a WE based on the hardware offloaded asynchronous copy instruction set; and write the WE into a second register, where the second register is located in the RNIC 120 and is communicable with the second engine 121.

[0156] The WE can include at least one of remote information Remote info, local information Local info, an operation code Opcode, a QPN (Queue Pair Number), and a tag TAG, where the TAG is used to mark a thread identifier ID or a block ID of a source of the RDMA operation request. The WE information collectively constitutes basic parameters of the RDMA data transmission operation, which can ensure efficient and secure execution of the RDMA data transmission.

[0157] Optionally, the first engine 131 can be configured to write the WE into the second register in a device direct PCIe P2P manner. The PCIe P2P manner is a hardware level direct communication, which bypasses the processor and the memory; and realizes asynchronous operation optimization, where after the WE is written into the second register, a hardware engine (the second engine) in communication with the second register can automatically process the task (e.g., store the WE into the WQE Buffer) without waiting for a software instruction; in addition, the direct hardware path can reduce data transmission delay.

[0158] In some embodiments of the present disclosure, the second engine 121 can be configured to parse the WE and store the WE in the WQE Buffer 123; and based on an occupancy rate of the WQE Buffer 123 being greater than or equal to a preset threshold, transmit a plurality of WE stored in the WQE Buffer 123 to a host memory of a network device.

[0159] In particular, the second engine 121 can receive a WE, for example, from a PCIe bus. After receiving the WE, the second engine 121 can verify the integrity of the WE information. The received WE information can be checked for completeness, data loss or damage. For example, the length in the WE information can be verified to be as expected, the checksum in the WE information can be verified to be correct, and the like. If the WE information is incomplete or has errors, the WE can be discarded and the error information can be recorded. In addition, the second engine 121 can also parse the WE and parse each field in the WE to determine the specific requirements of the operation. According to the information obtained by parsing, the second engine 121 can also allocate a cache space in the WQE Buffer 123 for the WE. If the on-chip WQE Buffer 123 is full, the WQE can be pushed to the host memory of the network device. This hierarchical storage can ensure performance through on-chip cache (for example, WQE Buffer 123), provide capacity redundancy through external memory, and ultimately achieve an efficient and reliable WQE processing pipeline.

[0160] Optionally, after parsing the WE, the second engine 121 can also schedule the processing order. For example, according to the priority of the operation, the queue state and other factors, the processing order of the WE can be determined. High-priority WE is processed first. This hardware arbitration mechanism automatically assigns task priorities, which can avoid queue blocking.

[0161] In addition, the second engine 121 can also generate a DB, wherein the DB can include at least one of queue pair information to be processed, a starting address of a plurality of WE, and a number of the plurality of WE; and transmit the DB to the RDMA engine, which can trigger the RDMA engine 122 to perform data processing, wherein the RDMA engine 122 can read the WE from the WQE Buffer 123 or the WQE external buffer in the above-mentioned host memory.

[0162] The generation and transmission of the DB itself is a sign of the start of the task. When the DB is written to a specific memory area (such as a conventional position of on-chip cache or host memory), the RDMA engine 122 can automatically start the processing flow after detecting the existence of the DB through hardware logic, without the intervention of software or additional interrupt signals. In addition, the DB can be located in a shared area of the on-chip cache or the host memory, which can be directly accessed by the RDM engine 122, which avoids data copying overhead. After completing the current data processing, the RDM engine 122 can immediately obtain the next data processing through the doorbell mechanism of the DB, without waiting for an interrupt response.

[0163] Optionally, the RDMA engine 122 can implement high-speed data transfer between hosts without involving kernel intervention of the operating system through direct memory access technology, thus having very low latency and high bandwidth. For example, the send WE contains a source address and a target address, and the RDMA engine 122 can directly move data from the source address to the send buffer of the RNIC 120 through direct memory access technology without copying.

[0164] In addition, the RDMA engine 122 can implement multi-type transmission support, such as Send / Receive, Read / Write, etc., where Send / Receive can be understood as bidirectional data transfer between local or remote nodes, and Read / Write can be understood as direct access to the memory of a remote node. This can eliminate context switching and interrupt overhead involving the xPU 130 and reduce data transmission delay to the order of microseconds. The RDMA engine 122 can support direct memory access across nodes through memory window technology, which can map the virtual address of a remote node to a physical address. In addition, each QP can contain an independent send queue and a receive queue, so the RDMA engine 122 supports concurrent processing of multiple task streams.

[0165] After the RDMA engine 122 completes the above data processing, the RDMA engine 122 can generate a CE message, and the second engine 121 receives the CE message and writes the received CE message to the first register, where the first register is located in the xPU 130 and can communicate with the first engine 131.

[0166] Both the first register and the second register described above can be understood as part of a BAR (Base Address Register) register. The first engine 131, the second engine 121, and the RDMA engine 122 can be located in different hardware modules or processing units in the RDMA data transmission network device and can communicate with each other through a NOC or shared registers (e.g., BAR registers). This avoids the first engine 131, the second engine 121, and the RDMA engine 122 directly transferring data, which can require complex hardware connections, reduces the complexity of design, and can ensure that different hardware modules or processing units work independently and reduce mutual interference.

[0167] In addition, the BAR register is a register reported by a PCIe device (e.g., RNIC, xPU, etc.) to a server during an enumeration phase, which is used to declare the memory space (e.g., configuration space, I / O space, memory mapping space, etc.) required by the device. The server can allocate a physical address for the BAR register, and thus the BAR register provides a standardized interface as part of memory-mapped I / O, which can complete data interaction through simple read / write operations, reducing design difficulty and power consumption.

[0168] Optionally, the second engine 121 can write the received CE message to the first register via PCIe P2P. This transmission method significantly improves the processing speed of RDMA data transfer tasks and the overall system performance. For example, PCIe P2P is a direct hardware-level communication method that bypasses the processor and memory; it achieves asynchronous operation optimization. After the CE is written to the first register, the hardware engine (first engine 131) communicating with the first register can automatically process the task (e.g., parse the CE) without waiting for software instructions; in addition, the direct hardware path can reduce data transfer latency.

[0169] Alternatively, the CE message may include at least one of the following: opcode, queue pair number QPN, and tag, where the TAG is used to identify the thread identifier ID of the RDMA operation request or the block ID of the source. Together, the CE information constitutes the basic parameters of the RDMA data transfer operation, ensuring efficient and secure execution of RDMA data transfer.

[0170] After receiving the CE message, the first engine 131 can generate a task completion message based on the CE message and transmit it to the NOC, which is then transmitted to the xPU that issued the RDMA operation request.

[0171] Therefore, in the RDMA data transfer scheme provided by the embodiments of this disclosure, the first engine is directly integrated inside the xPU, and the communication protocol including the hardware offload instruction set can be triggered by this hardware engine; the second engine is a dedicated preprocessing hardware engine, which is integrated in the RNIC, and can process DWQE at hardware speed, thereby freeing up the main processing unit resources of the CPU and RNIC. This reduces the number of data transfers during RDMA data transfer and reduces data latency; in addition, by directly offloading the RDMA data transfer task to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transfer task is decoupled from CPU / GPU resources, thus providing a high-efficiency and low-latency RDMA data transfer solution that does not rely on CPU / GPU resources for data transfer-intensive scenarios such as the MoE model.

[0172] Refer again Figure 1 , Figure 4 and Figure 6 As an implementation of the aforementioned RDMA data transmission method, the embodiments of this disclosure provide an RDMA data transmission network system, the system architecture 100 of which can be connected with... Figure 5 Corresponding to the method shown, this network system can be specifically applied to various electronic devices.

[0173] The RDMA data transmission network system architecture 100 provided by the embodiments of the present disclosure can include a plurality of network devices, for example, a first network device and a second network device, wherein the first network device can include an xPUA and an RNIC, and the second network device can include a server B and an xPUB. The xPUA and the xPUB both include a first engine, and the RNIC includes a second engine in communication with the first engine, a work queue buffer (WQE Buffer), and an RDMA engine in communication with the second engine.

[0174] The system architecture 100 can further include other network devices in addition to the first network device and the second network device, Figure 4 and Figure 6 The network devices in the system architecture 100 are shown by way of example in a small number, and a person skilled in the art can set the number, type, and connection relationship of the network devices in the system architecture 100 according to actual needs, which are not limited in the present disclosure.

[0175] In addition, in each of the plurality of network devices, the xPU can further include a core computing unit and a memory management unit, wherein the types of the core computing unit and the memory management unit of different types of xPU can be different. For example, in the case where the xPU includes a GPU, the core computing unit thereof can include an SM; in the case where the xPU includes a CPU, the core computing unit thereof can include an arithmetic logic unit, a control unit, etc.

[0176] In addition, when the xPU executes a program, the core computing unit generates a virtual address, and the memory management unit can be responsible for converting the virtual address into a physical address. Alternatively, in the case where the xPU includes a GPU or a CPU, the memory management unit thereof can include an MMU.

[0177] In each of the plurality of network devices, the RDMA engine of the RNIC can perform a plurality of data processing, wherein the data processing can include data transmission, memory access, and queue management, etc.

[0178] It should be noted that, since the content described in the foregoing RDMA data transmission network can be completely or partially applicable to the RDMA data transmission system described in the following, the related or similar content thereof will not be described again. However, a person skilled in the art can understand that the structural features, method processes, implementation principles, and technical effects achieved by the RDMA data transmission schemes described in the two are similar.

[0179] The first engine of the first network device can be configured to assemble a work element WE based on the hardware offloaded asynchronous copy instruction set, and transmit the WE to the second engine of the first network device; the second engine of the first network device can be configured to store the WE in the WQE Buffer, and transmit the WE to the RDMA engine of the first network device; the RDMA engine of the first network device can be configured to initiate a request to the second network device based on the WE; the second network device is configured to complete the request and feed back to the first engine of the first network device, wherein the first network device and the second network device are different network devices in the plurality of network devices in the system architecture 100.

[0180] As shown in Figure 1 The server 110 can be configured to create an interface corresponding to at least part of the MR of the RDMA, the interface information including a unique identifier UniqueID; based on the interface, allocate the MR to a plurality of network devices, and establish a MR page table of the network device, wherein the access memory credentials mKey of the plurality of MR page tables are all UniqueID; the plurality of network devices implement memory handle exchange; bind the obtained memory handle after exchange to the information of the WE, and establish a virtual address mapping relationship, so as to initiate remote access through the memory handle and UniqueID.

[0181] The UniqueID is used to identify the ID of the collective communication, and the synchronization of all nodes in the cluster has been completed at the time of communication initialization. In other words, at the starting point of the RDMA data operation, by starting the kernel program and carrying the UniqueID, it can be ensured that the operation can be accurately associated to a specific communication task, avoiding confusion between different communication tasks.

[0182] Optionally, the hardware offloaded asynchronous copy instruction set can include ibcp.async(dst.addr, src.addr, length), when Dst.addr is xPUB Reseverd VA, it means writing (Write) the memory pointed to by the xPUA local va of the local first network device to the memory of the remote second network device; when Src.addr is xPUB Reseverd VA, it means reading (Read) the memory of the remote second network device xPUB Memory to the memory pointed to by the xPUA local va of the local first network device.

[0183] In some embodiments of the present disclosure, the RDMA engine of the first network device is configured to obtain, based on the WE, a QPC corresponding to a queue pair (QPN) of the WE, and determine an operation type according to the QPC; the RDMA engine of the first network device is further configured to obtain, based on the WE, a memory handle and a UniqueID of the second network device; the RDMA of the first network device completes packet encapsulation based on the WE, and sends a request to the second network device.

[0184] After receiving the above request packet (e.g., a data write packet), the RDMA engine of the second network device can write payload data (Payload) in the request packet into the memory of the second network device according to the remote endpoint transport header, the UniqueID, and the length information, where the remote endpoint transport header can include a virtual address (xPUB VA, xPUB Virtual Address) of the second network device. By offloading the asynchronous copy instruction set and WE information through hardware, the second network device can be accurately located, and the xPU of the second network device is bypassed, reducing the intervention and overhead of the xPU, thereby significantly improving the efficiency of RDMA data transmission and avoiding the xPU from becoming a bottleneck of data transmission.

[0185] The RDMA engine of the second network device can also return an ACK (Acknowledgment, acknowledgment packet) to the RDMA engine of the first network device, to inform the first network device that the RDMA write request packet sent by the first network device has been successfully received and processed. This can ensure the reliability of RDMA data transmission.

[0186] After receiving the ACK, the RDMA engine of the first network device can generate a CE message according to the QP and the WQE, and deliver the CE message to the second engine. The second engine can assemble the received CE message, for example, to make the format and content of the message more standardized and unified. Then, it is delivered to the first engine in a P2P manner. By generating and delivering the CE message, the RDMA engine of the first network device can timely inform the hardware engine that the operation has been completed. This enables the hardware engine to timely perform subsequent processing, such as updating state information, releasing resources, or triggering new tasks. In addition, the P2P manner avoids the delay and bandwidth competition problems caused by message transmission through intermediate nodes or shared buses. Through the P2P manner, the CE message can be quickly delivered from the second engine to the first engine, improving the response speed and overall performance of the network device.

[0187] Therefore, in the RDMA data transfer scheme provided by the embodiments of this disclosure, the first engine is directly integrated inside the xPU, and the communication protocol including the hardware offload instruction set can be triggered by this hardware engine; the second engine is a dedicated preprocessing hardware engine, which is integrated in the RNIC, and can process DWQE at hardware speed, thereby freeing up the main processing unit resources of the CPU and RNIC. This reduces the number of data transfers during RDMA data transfer and reduces data latency; in addition, by directly offloading the RDMA data transfer task to the dedicated hardware engine through the hardware asynchronous instruction set, the RDMA data transfer task is decoupled from CPU / GPU resources, thus providing a high-efficiency and low-latency RDMA data transfer solution that does not rely on CPU / GPU resources for data transfer-intensive scenarios such as the MoE model.

[0188] According to embodiments of the present disclosure, the present disclosure also provides an electronic device, the electronic device comprising: at least one processor; and a memory communicatively connected to the at least one processor; wherein the memory stores instructions executable by the at least one processor, the instructions being executed by the at least one processor to enable the at least one processor to implement the RDMA data transmission method described in any of the above embodiments.

[0189] According to embodiments of this disclosure, this disclosure also provides a readable storage medium storing computer instructions that enable a computer to implement the RDMA data transmission method described in any of the above embodiments when executed.

[0190] According to embodiments of this disclosure, this disclosure also provides a computer program product that, when executed by a processor, can implement the RDMA data transmission method described in any of the above embodiments.

[0191] Figure 7 A schematic block diagram of an example electronic device 400 that can be used to implement embodiments of the present disclosure is shown. The electronic device is intended to represent various forms of digital computers, such as laptop computers, desktop computers, workstations, personal digital assistants, servers, blade servers, mainframe computers, and other suitable computers. The electronic device may also represent various forms of mobile devices, such as personal digital processors, cellular phones, smartphones, wearable devices, and other similar computing devices. The components shown herein, their connections and relationships, and their functions are merely illustrative and are not intended to limit the implementation of the present disclosure described and / or claimed herein.

[0192] like Figure 7As shown, the device 400 includes a computing unit 401 that can perform various appropriate actions and processes in accordance with a computer program stored in a read-only memory (ROM) 402 or a computer program loaded from a storage unit 408 into a random access memory (RAM) 403. In the RAM 403, various programs and data required for the operation of the device 400 can also be stored. The computing unit 401, the ROM 402, and the RAM 403 are connected to each other through a bus 404. An input / output (I / O) interface 405 is also connected to the bus 404.

[0193] A plurality of components in the device 400 are connected to the I / O interface 405, including: an input unit 406, such as a keyboard, a mouse, etc.; an output unit 407, such as various types of displays, speakers, etc.; a storage unit 408, such as a magnetic disk, an optical disk, etc.; and a communication unit 409, such as a network card, a modem, a wireless communication transceiver, etc. The communication unit 409 allows the device 400 to exchange information / data with other devices through a computer network, such as the Internet, and / or various telecommunication networks.

[0194] The computing unit 401 can be various general and / or special purpose processing components with processing and computing capabilities. Some examples of the computing unit 401 include, but are not limited to, a central processing unit (CPU), a graphics processing unit (GPU), various special-purpose artificial intelligence (AI) computing chips, various computing units running machine learning model algorithms, a digital signal processor (DSP), and any appropriate processor, controller, microcontroller, etc. The computing unit 401 performs various methods and processes described above, such as the RDMA data transfer method. For example, in some embodiments, the RDMA data transfer method can be implemented as a computer software program that is tangibly embodied in a machine-readable medium, such as the storage unit 408. In some embodiments, part or all of the computer program can be loaded and / or installed on the device 400 via the ROM 402 and / or the communication unit 409. When the computer program is loaded into the RAM 403 and executed by the computing unit 401, one or more steps of the RDMA data transfer method described above can be performed. Alternatively, in other embodiments, the computing unit 401 can be configured to perform the RDMA data transfer method by any other appropriate means, such as by means of firmware.

[0195] The various embodiments of the systems and techniques described above can be implemented in digital electronic circuitry, integrated circuitry, a field programmable gate array (FPGA), an application specific integrated circuit (ASIC), a system on a chip (SOC), a load programmable logic device (CPLD), computer hardware, firmware, software, and / or combinations thereof. These various embodiments can include implementation in one or more computer programs that are executable and / or interpretable on a programmable system including at least one programmable processor, which can be special or general purpose, coupled to receive data and instructions from, and to transmit data and instructions to, a storage system, at least one input device, and at least one output device.

[0196] Program code for carrying out methods of the present disclosure can be written in any combination of one or more programming languages. The program code can be provided to a processor or controller of a general purpose computer, special purpose computer, or other programmable data processing apparatus to produce a machine, such that the program code, when executed by the processor or controller, produces the functions / operations specified in the flowcharts and / or the block diagrams. The program code can be executed entirely on a machine, partially on a machine, partially on a machine as a stand-alone software package, or entirely on a remote machine or server.

[0197] In the context of the present disclosure, a machine-readable medium can be a tangible medium that contains or stores a program for use by or in connection with an instruction execution system, apparatus, or device. The machine-readable medium can be a machine-readable signal medium or a machine-readable storage medium. A machine-readable medium can include but is not limited to an electronic, magnetic, optical, electromagnetic, infrared, or semiconductor system, apparatus, or device, or any suitable combination of the foregoing. More specific examples of the machine-readable storage medium will include one or more lines of electrical conductors, portable computer disks, hard disks, random access memory (RAM), read-only memory (ROM), erasable programmable read-only memory (EPROM or Flash memory), optical fibers, portable compact disc read-only memories (CD-ROMs), optical storage devices, magnetic storage devices, or any suitable combination of the foregoing.

[0198] To provide for interaction with a user, the systems and techniques described here can be implemented on a computer having a display device (e.g., a CRT (cathode ray tube) or LCD (liquid crystal display) monitor) for displaying information to the user and a keyboard and a pointing device (e.g., a mouse or a trackball) by which the user can provide input to the computer. Other kinds of devices can be used to provide for interaction with a user as well; for example, feedback provided to the user can be any form of sensory feedback (e.g., visual feedback, auditory feedback, or tactile feedback); and input from the user can be received in any form, including acoustic, speech, or tactile input.

[0199] The systems and techniques described here can be implemented in a computing system that includes a back end component (e.g., as a data server), or that includes a middleware component (e.g., an application server), or that includes a front end component (e.g., a user computer having a graphical user interface or a Web browser through which a user can interact with an implementation of the systems and techniques described here), or any combination of such back end, middleware, or front end components. The components of the system can be interconnected by any form or medium of digital data communication (e.g., a communication network). Examples of communication networks include a local area network (LAN), a wide area network (WAN), and the Internet.

[0200] The computer system can include clients and servers. A client and server are generally remote from each other and typically interact through a communication network. The relationship of client and server arises by virtue of computer programs running on the respective computers and having a client-server relationship to each other. A server can be a cloud server, also known as a cloud computing server or cloud host, which is a host product in the cloud computing service system to solve the defects of large management difficulty and weak business scalability in traditional physical host and virtual private server (VPS, Virtual Private Server) services.

[0201] According to the technical scheme of the embodiment of the present disclosure, after receiving a training task indicating training of a generative large language model, a computing power cluster composed of single-type chips used for training and a target-type chip specifically used by the computing power cluster are determined first, then the performance evaluation of the target-type chip under a preset performance category is determined, and the performance evaluation of the corresponding performance category is weighted by combining the weight distribution of each preset performance category under different alternative training strategies and using the weight distribution, and the comprehensive training evaluation corresponding to each alternative training strategy is calculated according to the weighted performance evaluation, and finally the appropriate target training strategy is selected according to the evaluation parameters of each comprehensive training evaluation. That is, by applying the technical scheme provided in the embodiment, the appropriate target training strategy can be determined without performing trial training according to each alternative training strategy, reducing the resource overhead and time overhead required for trial training, and the target training strategy used for training the generative large language model by the computing power cluster composed of the target-type chip can be determined at a lower cost and quickly, so as to complete the training task only by spending less time overhead and resource overhead, thereby indirectly improving the model training efficiency.

[0202] It should be understood that the various forms of flow shown above can be reordered, added, or deleted steps. For example, each step described in the present disclosure can be executed in parallel, sequentially, or in a different order, as long as the desired results of the technical scheme of the present disclosure can be achieved, which is not limited herein.

[0203] The above specific embodiments do not constitute a limitation on the protection scope of the present disclosure. Those skilled in the art should understand that various modifications, combinations, sub-combinations and substitutions can be made according to design requirements and other factors. Any modifications, equivalent replacements and improvements made within the spirit and principles of the present disclosure shall be included in the protection scope of the present disclosure.

Claims

1. A method of RDMA data transfer, the method applied to a network device comprising a processor xPU and a RDMA network interface controller RNIC, the xPU comprising a first engine, the RNIC comprising a second engine in communication with the first engine, a work queue buffer WQE Buffer, and a RDMA engine in communication with the second engine, the method comprising: the first engine assembling a work element WE based on a hardware offload asynchronous copy instruction set, and transmitting the WE to the second engine; the second engine storing the WE in the WQE Buffer, and transmitting to the RDMA engine; the RDMA engine performing data processing based on the WE, the data processing comprising data transfer, memory access, and queue management; the first engine receiving feedback of the data processing transmitted via the second engine, wherein the hardware offload asynchronous copy instruction set comprises a multi-level synchronization control data transfer instruction, the multi-level synchronization control data transfer instruction comprises a thread-level synchronization control data transfer instruction, a thread group-level synchronization control data transfer instruction, a memory block-level synchronization control data transfer instruction, and a global-level synchronization control data transfer instruction.

2. The method of claim 1, wherein, the xPU further comprises a first register in communication with the first engine, the first engine receiving feedback of the data processing transmitted via the second engine comprising: after completion of the data processing, the RDMA engine generating a completion event CE message; the second engine receiving the CE message, and writing the received CE message to the first register; wherein the writing of the received CE message to the first register is in a device direct PCIe P2P manner, the CE message comprising at least one of an opcode Opcode, a queue pair number QPN, and a tag TAG, the TAG used to mark a thread identifier ID or a block ID of a source of the RDMA operation request.

3. The method of claim 2, wherein, further comprising: the first engine transmitting the feedback of the data processing to an xPU issuing the RDMA operation request through a network on chip NOC.

4. The method of claim 1, wherein, the RNIC further comprises a second register in communication with the second engine, the first engine assembling a work element WE based on a hardware offload asynchronous copy instruction set comprising: receiving an RDMA operation request; assembling the RDMA operation request into the WE based on the hardware offload asynchronous copy instruction set; writing the WE to the second register; wherein the writing of the WE to the second register is in a device direct PCIe P2P manner, the WE comprising at least one of a remote information Remote info, a local information Local info, an opcode Opcode, a queue pair number QPN, and a tag TAG, the TAG used to mark a thread identifier ID or a block ID of a source of the RDMA operation request.

5. The method of claim 1, wherein, the second engine storing the WE in the WQE Buffer, and sending to the RDMA engine comprising: parsing the WE and storing the WE in the WQE Buffer; transmitting a plurality of the WE stored in the WQE Buffer to a host memory of the network device based on an occupancy of the WQE Buffer being greater than or equal to a preset threshold.

6. The method of claim 5, wherein, The second engine storing the WE in the WQE Buffer and sending to the RDMA engine further includes: generating a doorbell signal DB, wherein the DB includes at least one of queue pair information to be processed, a starting address of a plurality of the WE, and a number of the plurality of the WE; transmitting the DB to the RDMA engine to trigger the RDMA engine to perform the data processing, wherein the RDMA engine reads the WE from the WQE Buffer or the host memory.

7. An RDMA data transmission method, applied to a network system including a plurality of network devices, the network device including a processor xPU and an RDMA network interface controller RNIC, the xPU including a first engine, the RNIC including a second engine in communication with the first engine, a work queue cache area WQE Buffer, and an RDMA engine in communication with the second engine, the method comprising: the first engine of a first network device assembling a work element WE based on a hardware offload asynchronous copy instruction set and transmitting the WE to the second engine of the first network device; the second engine of the first network device storing the WE in the WQE Buffer and transmitting to the RDMA engine of the first network device; the RDMA engine of the first network device initiating a request to a second network device based on the WE; the second network device completing the request and feeding back to the first engine of the first network device, wherein the first network device and the second network device are different network devices in the plurality of network devices; and the hardware offload asynchronous copy instruction set includes a multi-level synchronization control data transmission instruction, and the multi-level synchronization control data transmission instruction includes a thread-level synchronization control data transmission instruction, a thread group-level synchronization control data transmission instruction, a storage block-level synchronization control data transmission instruction, and a global-level synchronization control data transmission instruction.

8. The method of claim 7, further comprising: creating an interface corresponding to at least a portion of a memory region MR of the RDMA, the interface information including a unique identifier UniqueID; allocating the MR to a plurality of the network devices based on the interface and establishing MR page tables of the network devices, wherein a credential mKey of a plurality of the MR page tables accessing memory is the UniqueID; implementing memory handle exchange by a plurality of the network devices; binding the memory handle obtained after exchange to information of the WE and establishing a virtual address mapping relationship, so as to initiate remote access through the memory handle and the UniqueID.

7. An RDMA data transmission method, applied to a network system including a plurality of network devices, the network device including a processor xPU and an RDMA network interface controller RNIC, the xPU including a first engine, the RNIC including a second engine in communication with the first engine, a work queue cache area WQE Buffer, and an RDMA engine in communication with the second engine, the method comprising: the first engine of a first network device assembling a work element WE based on a hardware offload asynchronous copy instruction set and transmitting the WE to the second engine of the first network device; the second engine of the first network device storing the WE in the WQE Buffer and transmitting to the RDMA engine of the first network device; the RDMA engine of the first network device initiating a request to a second network device based on the WE; the second network device completing the request and feeding back to the first engine of the first network device, wherein the first network device and the second network device are different network devices in the plurality of network devices; and the hardware offload asynchronous copy instruction set includes a multi-level synchronization control data transmission instruction, and the multi-level synchronization control data transmission instruction includes a thread-level synchronization control data transmission instruction, a thread group-level synchronization control data transmission instruction, a storage block-level synchronization control data transmission instruction, and a global-level synchronization control data transmission instruction.

8. The method of claim 7, further comprising: creating an interface corresponding to at least a portion of a memory region MR of the RDMA, the interface information including a unique identifier UniqueID; allocating the MR to a plurality of the network devices based on the interface and establishing MR page tables of the network devices, wherein a credential mKey of a plurality of the MR page tables accessing memory is the UniqueID; implementing memory handle exchange by a plurality of the network devices; binding the memory handle obtained after exchange to information of the WE and establishing a virtual address mapping relationship, so as to initiate remote access through the memory handle and the UniqueID.

9. The method of claim 8, wherein, the RDMA engine of the first network device initiates a request to a second network device based on the WE, including: the RDMA engine of the first network device obtains a queue pair context (QPC) corresponding to a queue pair number (QPN) of the WE based on the WE, and determines an operation type according to the QPC; the RDMA engine of the first network device obtains the memory handle and the UniqueID of the second network device based on the WE; the RDMA of the first network device completes packet encapsulation based on the WE, and sends a request to the second network device.

10. An RDMA data transmission network device, comprising: a processor (xPU) including a first engine; an RDMA network interface controller (RNIC) including a second engine in communication with the first engine, a work queue buffer (WQE Buffer), and an RDMA engine in communication with the second engine, wherein the first engine is configured to assemble a work element (WE) based on a hardware offload asynchronous copy instruction set, and transmit the WE to the second engine; the second engine is configured to store the WE in the WQE Buffer, and transmit to the RDMA engine; the RDMA engine is configured to perform data processing based on the WE, and transmit feedback of the data processing to the first engine via the second engine, the data processing including data transmission, memory access, and queue management, wherein the hardware offload asynchronous copy instruction set includes a multi-level synchronization control data transmission instruction, the multi-level synchronization control data transmission instruction including a thread-level synchronization control data transmission instruction, a thread group-level synchronization control data transmission instruction, a memory block-level synchronization control data transmission instruction, and a global-level synchronization control data transmission instruction.

11. An RDMA data transmission network system, comprising a plurality of network devices, the network devices including a network device of a processor (xPU) and an RDMA network interface controller (RNIC), the xPU including a first engine, and the RNIC including a second engine in communication with the first engine, a work queue buffer (WQE Buffer), and an RDMA engine in communication with the second engine, wherein the first engine of a first network device is configured to assemble a work element (WE) based on a hardware offload asynchronous copy instruction set, and transmit the WE to the second engine of the first network device; the second engine of the first network device is configured to store the WE in the WQE Buffer, and transmit to the RDMA engine of the first network device; the RDMA engine of the first network device initiates a request to a second network device based on the WE; the second network device is configured to complete the request, and feed back to the first engine of the first network device, wherein the first network device and the second network device are different network devices in the plurality of network devices; and The hardware offloaded asynchronous copy instruction set includes multi-level synchronization control data transfer instructions, and the multi-level synchronization control data transfer instructions include thread-level synchronization control data transfer instructions, thread group-level synchronization control data transfer instructions, memory block-level synchronization control data transfer instructions, and global-level synchronization control data transfer instructions. 12.An electronic device comprising: at least one processor; and a memory communicatively connected to the at least one processor; wherein the memory stores instructions executable by the at least one processor, and the instructions are executed by the at least one processor to enable the at least one processor to perform the RDMA data transfer method of any one of claims 1 to 9. 13.A non-transitory computer-readable storage medium storing computer instructions for causing a computer to perform the RDMA data transfer method of any one of claims 1 to 9. 14.A computer program product comprising a computer program which, when executed by a processor, implements the steps of the RDMA data transfer method according to any one of claims 1 to 9.

Images