A training phase perception type SHARP intra-network set communication operator dynamic arrangement system and method

Abstract

The present application relates to a kind of training phase perception type SHARP in-network collection communication operator dynamic scheduling system and method, system includes: training framework domain: perception training phase and generate standardization phase descriptor;Host channel adapter domain: descriptor is forwarded to switch SMA by out-band channel, and listens to GPU progress token to realize submission gate;Switch CFU dynamic scheduling domain: according to strategy table, reconfigure parameter in CFU pipeline idle gap, realize no-interruption dynamic scheduling.The present application establishes a light training phase perception collaborative link between training framework in host side and CFU in switch side, realizes the real-time dynamic scheduling of CFU execution strategy with training phase, so that the in-network computing acceleration benefit of SHARP is extended from the current limited All-Reduce scene to the complete communication life cycle of AI large model training.

Description

Technical Field

[0001] This invention belongs to the field of domain network interconnection and artificial intelligence technology, specifically relating to a dynamic orchestration system and method for perceptual SHARP network set communication operators during the training phase. Background Technology

[0002] In large-scale distributed AI training scenarios, computing nodes (typically servers equipped with GPUs / NPUs) are interconnected via InfiniBand networks and frequently perform collective communication operations, including All-Reduce, Reduce-Scatter, and All-Gather, to synchronize gradients or parameters across all nodes. Research indicates that in training clusters ranging from kilocalories to tens of thousands of kilocalories, collective communication accounts for 30% to 50% of the total training time and is one of the core bottlenecks to overall performance.

[0003] NVIDIA's Scalable Hierarchical Aggregation and Reduction Protocol (SHARP) reduces the amount of data that previously required multiple rounds of transmission between all GPUs from O(N) to O(1) by directly performing reduction operations on a dedicated Collective Functional Unit (CFU) integrated within the InfiniBand switch. This significantly reduces network traffic and latency in aggregated communication. SHARPv4 is integrated into the NVIDIA Quantum-X800 XDR 800G InfiniBand switching platform, providing up to 14.4 TFLOPS of intranet computing power and adding support for FP8 precision and MPI_Alltoall. However, existing SHARP and similar intranet aggregated acceleration technologies suffer from the following key technical defects, which limit their actual acceleration effect in AI large-scale model training scenarios: Defect 1: The set operator type is statically bound throughout training, making it impossible to perceive changes in communication characteristics during the training phase. Distributed large model training has a highly regular three-stage structure: Forward Pass stage, All-Gather communication of parameters is generated in Fully Sharded Data Parallelism (FSDP) or Mixed Precision Pipeline Parallelism to reconstruct the complete weights of each layer; Backward Pass stage, each node performs Reduce-Scatter on its local gradients, distributing gradient shards to the corresponding parameter nodes; Optimizer Step stage, some frameworks (such as DDP) perform All-Reduce to aggregate the gradients and broadcast them back to all nodes, or the ZeRO series optimizer performs All-Gather of sharded parameters to complete parameter synchronization.

[0004] In existing SHARP implementations, the aggregation tree and aggregation group are statically configured by the SHARP Aggregation Manager (AM) during training task initialization. The type of set operator executed by the CFU is bound to the aggregation group creation phase. Throughout the training lifecycle, this leads to two types of resource waste: First, during the FSDP backpropagation phase, the communication semantics are Reduce-Scatter, while the statically bound All-Reduce will perform a complete reduction within the network before being broadcast by the result root node, generating additional, unnecessary broadcast traffic in the FSDP scenario, accounting for approximately N times the redundancy of Reduce-Scatter communication (N being the number of parallel nodes). Second, during the forward propagation All-Gather phase, existing SHARP does not perform intra-network acceleration for All-Gather (because All-Gather is essentially a broadcast operation, and the CFU reduction logic is ineffective for it), resulting in CFU resources being completely idle during this phase.

[0005] Defect 2: The CFU numerical precision configuration is a globally static parameter, which does not match the precision requirements of each training stage. Modern large models trained with FP8 mixed precision (such as the FP8 path supported by NVIDIA Transformer Engine) perform matrix multiplication in FP8 during forward propagation, while gradient accumulation in backpropagation is usually performed in BF16 or FP32 to ensure convergence stability. The existing SHARP CFU precision mode is statically configured when the aggregation group is created (e.g., FP32 throughout), and cannot switch precision within iterations. This results in CFU performing intra-network reduction with excessively high precision in FP8 forward propagation scenarios, consuming unnecessary CFU computational resources; while in BF16 gradient backpropagation scenarios, using FP8 precision configuration introduces unacceptable gradient quantization errors.

[0006] Defect 3: The submission timing of SHARP aggregation groups is unaware of the GPU computation stream progress, leading to insufficient utilization of the communication-computation overlap window. Modern distributed training frameworks (such as Megatron-LM, DeepSpeed, and PyTorch FSDP) widely adopt the communication-computation overlap strategy: when the GPU is performing backpropagation at layer L, the gradient reduce-scatter of layer L+1 is performed asynchronously in the background. However, the activation operation of SHARP aggregation groups (i.e., submission to CFU for execution) depends on the initiation time of the NCCL call on the host side. The NCCL itself does not accurately perceive the completion progress of the GPU stream (CUDAStream). When the GPU computation stream jitters, the NCCL may prematurely submit gradient data that is not yet ready to the SHARP aggregation group, causing the CFU to wait for the data to arrive and occupying aggregation tree node resources; or delayed submission may cause the GPU computation stream to idle while waiting for communication to complete, creating a pipeline bubble.

[0007] Defect 4: The current SHARP does not support the phased intra-network acceleration combination of Reduce-Scatter and All-Gather in FSDP scenarios. In the fully sharded parallel mode such as ZeRO-3 / FSDP, the standard gradient synchronization process is divided into two steps: First, each node executes Reduce-Scatter to aggregate the gradient shards it is responsible for reducing; second, each node executes All-Gather to broadcast the global parameters to all nodes. These two steps must be sequential in timing, but they belong to different communication primitives. The current SHARP only executes one operator (such as All-Reduce) for a single aggregation group call, and cannot directly connect the intra-network reduction result of Reduce-Scatter to the All-Gather acceleration of the next aggregation group. It needs to be relayed on the GPU side, introducing additional device memory read / write and host-side scheduling overhead.

[0008] The aforementioned defects collectively result in the actual utilization rate and acceleration benefits of SHARP network computing resources being far lower than its theoretical peak hardware capacity in large-scale AI model training scenarios with tens of thousands of kilowatts. Summary of the Invention

[0009] One objective of this invention is to provide a dynamic orchestration system for SHARP network aggregate communication operators during the training phase.

[0010] To achieve the above objectives, the technical solution adopted by the present invention is as follows: A training-phase perceptual SHARP network intra-network ensemble communication operator dynamic orchestration system, comprising: Training framework domain: Running on the host side of the GPU computing node, it is responsible for sensing and extracting the current stage state of the training framework, generating standardized training stage descriptors, and transmitting stage semantics to the switch side through the host channel adapter. Host Channel Adapter Domain: Running at the network interface card firmware layer, it is responsible for receiving stage descriptors from the training framework domain. Before the aggregated communication operation is submitted to the SHARP aggregation group, it forwards the stage descriptors to the subnet management agent of the target switch through the out-of-band control channel of the management datagram. At the same time, it is responsible for listening to the GPU computing flow progress token to achieve precise gating of the communication submission timing. Switch CFU Dynamic Orchestration Domain: Running in the aggregation node management processor inside the switch, it is responsible for receiving stage descriptors from the host channel adapter domain, reconfiguring CFU execution parameters at runtime according to the stage-operator mapping strategy table, and completing parameter switching during the idle gap of the CFU pipeline after the current aggregation operation is completed, so as to achieve uninterrupted dynamic orchestration.

[0011] Preferably, in the above technical solution, the training framework domain includes: Training framework: Used to perform AI training tasks, and perceives the stage state through gradient hooks and / or interception of the NCCL collection communication library; Stage-aware probe module: Connected to the training framework, it receives raw information from the training framework and performs feature extraction and standardization of training stage descriptors.

[0012] Preferably, in the above technical solution, the host channel adapter domain includes: PD Encoding and MAD Forwarding Module: Used to receive stage descriptors, encode them into InfiniBand management datagram attributes, and forward them to the subnet management agent of the target switch in advance, through the out-of-band control channel of the high-priority queue, before the submission of the aggregate communication operation. GPU compute stream progress gating module: Used to listen to the progress tokens of the GPU compute stream and generate gating ready signals based on the progress to control the timing of enqueuing of collection communication operations.

[0013] Preferably, in the above technical solution, the switch CFU dynamic orchestration domain includes: PD Reception and Policy Query Module: Used to receive and parse phase descriptors and validity verifications through the subnet management agent, and query the corresponding CFU parameter vector based on the phase-operator mapping policy table; CFU dynamic parameter reconfigurator: connected to the PD receiving and policy query module, used to detect the CFU pipeline idle interval, and write the queried CFU parameter vector into the CFU register group through atomic write operation to dynamically reconfigure the CFU execution parameters; The aggregated function unit execution engine is configured by the CFU dynamic parameter reconfigurator and is used to perform intranet reduction / aggregation calculations on the input aggregated data and output the results according to the configured route.

[0014] More preferably, the parameter vector of the stage-operator mapping strategy table includes operator type, precision mode, routing strategy, and chain enable.

[0015] More preferably, the operator types include All-Reduce, Reduce-Scatter, All-Gather, and Reduce-Scatter+All-Gather chained operations; the precision modes include FP32, BF16, FP16, and FP8; and the output routes include broadcast routes, fragmented routes, and direct-path routes.

[0016] Another objective of this invention is to provide a method for dynamically arranging aggregate communication operators within a SHARP network during the training phase.

[0017] To achieve the above objectives, the technical solution adopted by the present invention is as follows: A method for dynamic orchestration of ensemble communication operators within a SHARP network during the training phase includes the following steps: S1: Training Framework Stage Awareness and Training Stage Descriptor Construction: Awareness of the current stage of the training framework, forming a raw field set for the training stage descriptor; acquisition of GPU computation flow progress tokens and filling them into the corresponding fields of the stage descriptor; assembly of a 32-byte stage descriptor, and pushing it into the send queue on the host channel adapter side. S2: Stage descriptor encoding, transmission, and gating on the host channel adapter side: Read the stage descriptor from the send queue, encode and encapsulate it into a management datagram and send it to the switch subnet management agent, and write it to the stage descriptor receive buffer; poll the GPU progress, and when the progress reaches the target recorded in the stage descriptor, allow the set operation to commit. S3: Switch-side stage descriptor reception, policy query, and CFU parameter reconfiguration: Read the stage descriptor in the receive buffer, check its validity and the aggregation group it belongs to; use the fields in the stage descriptor as indexes to query the stage-operator mapping policy table, detect CFU pipeline idle gaps, and write the queried parameters into the CFU control register group. S4: Intra-network aggregated communication execution after CFU dynamic orchestration: Each node sends data fragments to the CFU, the CFU performs corresponding calculations and routing operations, and sends the results to the corresponding nodes. S5: Post-execution status feedback and strategy table update: CFU reports execution statistics and writes them to the statistics buffer; it uniformly polls and visualizes the statistical data, and adjusts and issues the updated strategy table accordingly.

[0018] Preferably, in the above technical solution, when sensing the current stage of the training framework: relevant fields are extracted by means of gradient hook interception and / or NCCL set communication group interception to form a set of original fields of the stage descriptor to be encoded.

[0019] Preferably, in the above technical solution, when collecting GPU computing flow progress tokens: before the NCCL collection communication operation is enqueued, a progress marker is inserted into the current GPU's command queue, and the sequence number of the marker is filled into the stage descriptor as a progress token.

[0020] Preferably, in the above technical solution, the stage descriptor read in S2 is the head stage descriptor of the transmission queue with a higher priority than the ordinary data transmission queue, and it is encoded as the attribute data field of the InfiniBand management data packet, and the management data packet is a no-reliable management data packet.

[0021] Preferably, when checking the validity of the stage descriptor and its associated aggregation group, the following steps are taken: check the Valid flag bit in the stage descriptor and verify whether the SHARP aggregation group identifier recorded therein belongs to the aggregation tree node range under the responsibility of this switch.

[0022] Preferably, in S3, when the CFU execution engine is detected to be in an idle gap where the previous aggregation operation has been completed and the data packet for the next aggregation operation has not yet arrived, the parameter vector obtained from the query is written to the CFU control register group by an atomic write operation to complete the parameter reconfiguration.

[0023] More preferably, if the operator type and precision mode indicated by the stage descriptor are the same as the current configuration in the CFU control register group, the reconfiguration step is skipped.

[0024] Preferably, in the above technical solution, when the CFU is configured to perform Reduce-Scatter intranet acceleration: the CFU performs element-wise accumulation reduction on the gradient shards sent by each computing node, and sends the reduction result to the corresponding parameter responsible node according to the sharding routing strategy.

[0025] Preferably, in the above technical solution, when the CFU is configured to perform chained intranet pipeline acceleration of Reduce-Scatter and All-Gather: the CFU sequentially executes Reduce-Scatter reduction and All-Gather broadcast, wherein the output of Reduce-Scatter is used as the input of All-Gather, and is executed in a continuous stream manner within the CFU pipeline, and intermediate results are not written back to GPU memory.

[0026] Preferably, in the above technical solution, when the CFU is configured to perform All-to-All intra-network acceleration: each node sends the routed Token feature vector to the CFU in fragments according to the destination Expert node, and the CFU performs non-standard routing forwarding operation to forward each fragment directly to the destination Expert node.

[0027] Preferably, the statistical information in S5 includes: the type of operator actually executed, the precision mode, the reduction time, the number of participating nodes, and the amount of data.

[0028] Preferably, in S5 of the above technical solution: the statistical data of each switch is polled and visualized through the InfiniBand subnet management interface; the configuration of the phase-operator mapping strategy table is adjusted according to the visualization results; and the data is distributed to all switches in the network through the batch update interface.

[0029] This invention breaks the static binding mode of training semantics between set operator type, numerical precision configuration and submission timing in the existing SHARP system. By establishing a lightweight training phase-aware collaborative link between the host-side training framework and the switch-side CFU, the CFU execution strategy can be dynamically arranged in real time according to the training phase. This extends the intra-network computing acceleration benefits of SHARP from the current limited All-Reduce scenario to the entire communication lifecycle of AI large model training.

[0030] Due to the application of the above technical solution, the present invention has the following advantages compared with the prior art: 1. Eliminates redundant broadcast traffic in FSDP scenarios, reducing network bandwidth usage by approximately N times: By switching the CFU operator from All-Reduce to Reduce-Scatter during the backpropagation phase, unnecessary global broadcast steps generated by All-Reduce are eliminated in FSDP training. The network traffic for gradient communication is reduced from twice that of full All-Reduce (one aggregation + one broadcast) to only one fragment reduction transmission, theoretically saving N / 2 times the bandwidth (N is the number of data parallel nodes). 2. Improved CFU utilization during the forward propagation phase of FP8, with coordinated optimization of accuracy and computing efficiency: By switching the CFU accuracy to FP8 during the forward propagation phase, the required accuracy for All-Gather / All-to-All operations is guaranteed (FP8 has no accuracy loss in routing operations), while the computational throughput of each CFU reduction is increased by about 2 to 4 times (compared to FP32). The reduction time is shorter with the same amount of data, freeing up more CFU resources for the subsequent backward propagation phase. 3. RS+AG chained intranet pipeline eliminates GPU-side memory read / write overhead: In the optimizer step, RS+AG chained execution allows the RS reduction output to be directly piped to the AG broadcast input within the CFU, eliminating one GPU memory write (RS result) and one GPU memory read (AG input) on the GPU side in the traditional step-by-step implementation. On the A100 GPU, the measured HBM memory read / write bandwidth is about 2TB / s, and the overhead of two memory accesses for 1GB parameters is about 1 millisecond. The chained solution eliminates this overhead, which can reduce the overall optimizer step communication time by about 5%~10%.

[0031] 4. Eliminating GPU pipeline bubbles caused by premature communication submissions: By precisely aligning the submission time of collection communication with the completion progress of GPU computation flow, the invalid occupation of CFU waiting for data before the GPU has completed gradient calculation and the idle bubbles of the GPU waiting for communication to complete are avoided. It is expected to increase the computation-communication overlap rate from about 70%~80% in the existing scheme to more than 90%. Attached Figure Description

[0032] Appendix Figure 1 This is a diagram of the dynamic orchestration system architecture for the perceptual SHARP network set communication operator during the training phase in this application. Appendix Figure 2 This is a detailed diagram illustrating the bit width of the 32-bit stage descriptor core control header in this application. Appendix Figure 3 This is a schematic diagram of a pipeline in the distributed execution mode of existing technology; Appendix Figure 4 This is a schematic diagram of the RS+AG chain network pipeline in this application; Appendix Figure 5 This is a timing diagram of the backpropagation phase of the Wanka LLaMA-3 training cluster in this embodiment. Detailed Implementation

[0033] The technical solution of the present invention will now be clearly and completely described with reference to the accompanying drawings. Obviously, the described embodiments are only some, not all, of the embodiments of the present invention. Based on the embodiments of the present invention, all other embodiments obtained by those skilled in the art without creative effort are within the scope of protection of the present invention.

[0034] like Figure 1 As shown: A training-phase perceptual SHARP network-based dynamic orchestration system for aggregated communication operators, comprising three functional domains working collaboratively: the training framework domain, the host channel adapter domain, and the switch CFU dynamic orchestration domain. The overall system architecture is as follows. Figure 1 As shown below, the three functional domains will be described in detail.

[0035] Training Framework Domain: Running on the host CPU of the GPU computing node, it is responsible for sensing and extracting the current stage state of the training framework, generating standardized training stage descriptors (PDs), and transmitting stage semantics to the switch side through the host channel adapter (HCA, ConnectX series network card).

[0036] In this embodiment, the training framework domain specifically includes the training framework and the phase sensing probe module (PSP).

[0037] The training framework is used to perform AI training tasks, such as the PyTorch framework. Its core components (gradient hooks, NCCL communication library) are the source of stage information. It does not directly participate in orchestration, but provides the system with the original stage state. In this embodiment, the training framework perceives the stage state through gradient hooks and / or intercepting the NCCL set communication library.

[0038] The stage-aware probe module is connected to the training framework to receive stage states from the training framework, perform feature extraction and standardize the training stage descriptor, and specifically generate a concise 32-byte stage descriptor.

[0039] The Host Channel Adapter Domain (HCA Domain) runs at the NIC firmware layer (ConnectX) and is responsible for receiving stage descriptors from the training framework domain. Before the aggregated communication operation is submitted to the SHARP aggregation group, the stage descriptor is forwarded to the subnet management agent (SMA) of the target switch through the out-of-band control channel of the InfiniBand Management Datagram (MAD). At the same time, it is responsible for listening to the GPU computing flow progress token to achieve precise gating of the communication submission timing.

[0040] In this embodiment, the host channel adapter domain specifically includes a PD encoding and MAD forwarding module and a GPU computing flow progress gating module (Progress Gate).

[0041] The PD encoding and MAD forwarding module is used to receive the stage descriptor, encode it into an InfiniBand management datagram (MAD) attribute, and forward it to the subnet management agent (SMA) of the target switch through an out-of-band control channel of a separate, high-priority queue, at a pre-set time (e.g., 1-2 μm) before the submission of the aggregate communication operation.

[0042] The GPU compute stream progress gating module is used to listen to the progress tokens of the GPU compute stream (CUDA Stream) and generate gating ready signals based on the progress to control the timing of enqueuing ensemble communication operations.

[0043] The Switch CFU Dynamic Orchestration Domain runs within the Aggregation Node (AN) management processor of the InfiniBand switch (Quantum-X800). It is responsible for receiving stage descriptors from the host channel adapter domain, reconfiguring the CFU execution parameters (operator type, precision mode, output routing policy) at runtime according to the stage-operator mapping policy table, and completing parameter switching during the idle gap of the CFU pipeline after the current aggregation operation is completed, thus achieving uninterrupted dynamic orchestration.

[0044] In this embodiment, the switch CFU dynamic orchestration domain includes a PD receiving and policy query module, a CFU dynamic parameter reconfigurator, and a collection function unit execution engine.

[0045] The PD receiving and policy query module is used to receive and parse phase descriptors and validity verifications through the Subnet Management Agent (SMA), and query the corresponding CFU parameter vectors based on the phase-operator mapping policy table.

[0046] The Stage-Operator Mapping Strategy Table (POMT) is the system's decision manual, a pre-built lookup table that stores the optimal CFU hardware configuration for different training stages, such as operator type, precision mode, routing strategy, and chain enable.

[0047] The CFU dynamic parameter reconfigurator is connected to the PD receiver and policy query module to detect the CFU pipeline idle gap. Within a nanosecond-level (less than 10 nanoseconds) time window, it writes the queried CFU parameter vector into the CFU register group through an atomic write operation to dynamically reconfigure the CFU execution parameters.

[0048] The execution parameters of CFU specifically include operator type, precision mode, and output route. Specifically: operator type includes All-Reduce, Reduce-Scatter, All-Gather, and Reduce-Scatter+All-Gather (RS+AG) chained operations; precision mode includes FP32, BF16, FP16, and FP8; output route includes broadcast route, fragmented route, and direct route.

[0049] The configuration of the aggregated functional unit execution engine is set by the CFU dynamic parameter reconfigurator, which is used to perform dynamic precision intra-network reduction / aggregation calculations on the aggregated data of the input aggregation tree AN nodes, and output the results as configured for broadcast / fragmentation routing.

[0050] The stage descriptor (PD) is the core data structure in this application. It is a new lightweight data structure introduced on the call path of the SHARP aggregation group activation interface (such as sharp_coll_do_all_reduce()) as a standardized contract between the training framework and the switch-side CFU dynamic orchestration controller.

[0051] PD uses a fixed-length compact binary format with a total size of 32 bytes. Its field definitions are shown in Table 1.

[0052] Table 1 The definitions of each bit in the Control Header field (32 bits in total) are shown in Table 2, and the detailed diagram of the width bit is shown below. Figure 2 As shown.

[0053] Table 2 A dynamic orchestration method for SHARP network set communication operators in the training phase is proposed. This method is implemented through the aforementioned system, and its complete data flow includes the following five main steps.

[0054] S1: Training framework phase awareness and training phase descriptor (PD) construction.

[0055] S1.1: Event interception by the phase awareness probe module: The Stage Aware Probe Module (PSP) senses the current stage of the training framework in the following two ways: Gradient Hook Interception: PSP registers the `post_accumulate_grad_hook` callback function with training frameworks such as PyTorch. When the training framework executes backpropagation and gradient accumulation, and triggers this hook, PSP extracts the gradient tensor dimension (Tensor_Element_Count) and data type (used to determine `Prec_Hint`) of the current layer, and sets the `Phase_ID` to 01 (BWD). NCCL group communication group interception: PSP inserts the phase annotation interface ncclSetPhaseHint(phase_id, op_hint, prec_hint) between the ncclGroupStart() and ncclGroupEnd() calls of NCCL. The host-side driver fills in the Phase_ID, Op_Hint, and Prec_Hint fields according to the logic currently being executed by the upper-layer AI framework (such as the Megatron-LM's allreduce_gradients() call being triggered after backpropagation and before the optimizer step), forming the set of PD raw fields to be encoded.

[0056] S1.2: GPU compute flow progress token collection: Before each NCCL set communication operation is enqueued, the PSP inserts a lightweight progress marker into the current CUDAStream through the GPUDirect Async interface, obtains the sequence number of the marker as the Stream_Progress_Token, and fills the corresponding field of the PD. This token will serve as the gating condition for the switch-side CFU to start execution.

[0057] S1.3: PD build complete: The PSP assembles all the collected fields into a 32-byte PD binary structure, and expresses the phase semantics of distributed AI training in a structured way through a fixed-length compact binary descriptor (32 bytes). Its key embedded phase semantic fields include Phase_ID, Op_Hint, Prec_Hint, and Parallel_Mode. This PD is pushed into the PD SendQueue (PDSQ) on the HCA side. PDSQ is a dedicated queue added in the HCA firmware layer, with a higher priority than the ordinary data transmission queue, to ensure that the PD completes pre-transmission before the InfiniBand data packet of the corresponding set communication operation arrives at the switch.

[0058] S2: HCA-side PD encoding, out-of-band transmission, and submission gating.

[0059] S2.1: PD Encoding and MAD Packaging: The PD encoding module of the HCA firmware layer reads the PD at the head of the PDSQ queue and encodes it into the attribute data field of the InfiniBand Management Datagram (MAD). It then constructs a management datagram with the destination address corresponding to the InfiniBand switch subnet management agent (SMA). The MAD is transmitted using Unreliable Management Datagram, with a latency of approximately 1 to 2 microseconds, which is lower than the end-to-end latency established by the SHARP data plane InfiniBand RC transmission, ensuring that the PD arrives at the switch earlier than the data.

[0060] S2.2: PD is pre-transmitted to the switch SMA via an out-of-band channel: HCA sends the encapsulated MAD to the target switch's SMA via the InfiniBand management port. After receiving it, the SMA writes the raw PD bytes into the PD Receive Buffer (PDRB) of the switch's AN management processor. This is then pre-transmitted to the switch side via the InfiniBand MAD out-of-band control channel before the arrival of the aggregated communication data packet. This establishes a direct correlation path between the training framework semantics and the CFU hardware execution strategy, unlike the existing static binding mode where the SHARP aggregation group configuration is completely independent of the upper-layer training semantics. The entire PD out-of-band transmission path is independent of the SHARP data plane RC transmission channel that carries the gradient data and does not occupy the data bandwidth of the aggregation tree.

[0061] S2.3: GPU compute flow progress gating: The Progress Gate Module (PGM) of the HCA firmware layer continuously polls the stream progress counter reported by the GPU through the GPUDirect Async interface. If the current completion progress number of the GPU computation stream is greater than or equal to the target progress number recorded in the Stream_Progress_Token field of the PD, the PGM sends a "ready" enable signal to the Collective Operation Submit Queue (COSQ) of the NCCL, allowing the SHARP aggregation group activation request for this collection operation to be formally enqueued and sent. If the Overlap_Gate bit is 0, this gating logic is skipped and submission is allowed directly.

[0062] By carrying a GPU CUDA Stream progress token (Stream_Progress_Token) in the Product Processor (PD), a gating mechanism is implemented at the HCA firmware layer to precisely align the set communication commit timing with the GPU computation stream completion progress. The HCA Progress Gating Module (PGM) blocks the SHARP aggregation group activation request of the set operation before the GPU progress counter reaches the target token value, reducing the time difference between the communication commit time and the GPU computation completion time to the HCA firmware polling cycle level (approximately 100 nanoseconds). This systematically eliminates CFU wait times caused by early communication commits and GPU pipeline bubbles caused by delayed communication commits.

[0063] S3: Switch-side PD reception, policy query, and CFU parameter reconfiguration.

[0064] S3.1: PD Reception and Validity Verification: The switch AN management processor reads the newly arrived PD from the PDRB, checks the Valid bit (Bit[0]) to confirm its validity, and checks the SHARP aggregation group ID recorded in the PD (determined by the mapping relationship between the MAD source port and the registered aggregation tree) to ensure that this PD belongs to the aggregation tree node range under the responsibility of this switch.

[0065] S3.2: Stage-operator mapping strategy table lookup: The AN management processor uses the Phase_ID (2 bits), Op_Hint (3 bits), Prec_Hint (3 bits), and Parallel_Mode (4 bits) fields in the PD to form a 12-bit index to query the Phase-Operator Mapping Table (POMT) preloaded in the switch's on-chip SRAM. The POMT is a lookup table with 2^12 = 4096 entries, and each entry contains the following CFU execution parameter vector as shown in Table 3.

[0066] Table 3 S3.3: Dynamic reconfiguration of CFU execution parameters: The AN management processor detects the pipeline status of the current CFU execution engine and determines whether it is in an IdleGap: that is, within the time window when the last batch of data packets of the previous aggregation operation has been reduced and output, and the data packets of the next aggregation operation have not yet started arriving. After confirming the IdleGap, the AN management processor writes the CFU execution parameter vector from the POMT query results to the CFU Control Register File (CCRF) using an atomic register write operation, completing the parameter reconfiguration. The atomic write operation is completed in a single clock cycle through the internal register bus, with a switching latency of no more than one CFU pipeline cycle (approximately 10 nanoseconds), without interrupting other ongoing aggregation operations. This differs from the static configuration limitation that existing CFU parameters can only be modified during the training initialization phase of aggregation group destruction / reconstruction. If the operator type and precision indicated by the newly arrived PD are exactly the same as the existing configuration in the current CCRF, the reconfiguration step is skipped, and the AN management processor directly notifies the POMT to increment the hit cache counter, saving unnecessary register write operations.

[0067] S4: Execution of intranet aggregate communication after CFU dynamic orchestration.

[0068] S4.1: Accelerated execution of Reduce-Scatter within the network (BWD stage): When Phase_ID=01 (BWD) and Op_Hint=001 (Reduce-Scatter), CFU is configured to perform Reduce-Scatter intranet acceleration: Each compute node's HCA sends its local gradient fragments (addressed by destination fragment index) to the AN node of the aggregation tree via the SHARP RC transmission channel; After receiving the corresponding fragments from all child nodes, the CFUs at each level of the AN node perform element-wise accumulation reduction in the configured precision mode (such as BF16). The reduction results are not broadcast (Output_Route=SR, fragmented routing). Instead, the AN root node directly sends the reduction results of each fragment along the downlink of the aggregation tree to the corresponding parameter responsible node through the fragmented routing strategy. The HCA of each parameter-responsible node writes the received gradient slices into the corresponding receive buffer in the GPU memory, completing the Reduce-Scatter operation without requiring the GPU to perform an additional ring-allreduce slice step.

[0069] S4.2: RS+AG chained intranet accelerated execution (FSDP OPT phase): When Phase_ID=10 (OPT) and RS_AG_Chain bit=1, CFU is configured to perform chained intranet pipeline acceleration of Reduce-Scatter and All-Gather: RS phase: Same as S4.1, the gradient fragments of each node are distributed to the corresponding nodes after being reduced within the CFU network; AG Phase: After each node receives the parameter fragments completed for the protocol it is responsible for, it directly sends them to the uplink of the aggregation tree through the All-Gather broadcast path of CFU (Output_Route=BR, broadcast); After the AN root node collects all parameter fragments, it broadcasts the complete parameter vector downlink to all compute nodes via the SHARP RC broadcast path; The entire RS+AG chain operation is executed in a continuous flow within the CFU pipeline. The output buffer of RS directly serves as the input buffer of AG, eliminating the need for GPU-side memory read / write operations as an intermediary and removing the overhead of GPU memory read / write operations and HCA-side communication task resubmission in the traditional two-step separation implementation.

[0070] This chained execution architecture, which pipelines the Reduce-Scatter reduction output buffer directly within the CFU to the All-Gather broadcast input, eliminates the data transfer overhead that must pass through GPU memory in the traditional two-step separation implementation in FSDP / ZeRO-3 scenarios, achieving end-to-end intranet pipelined processing from gradient reduction to parameter broadcasting. Figure 3 , 4 As shown: Compared with the traditional step-by-step execution mode, the chain pipeline in this embodiment completes the pass-through process inside the switch CFU through PD indication enable, completely avoiding the relay of GPU HBM video memory, thereby significantly reducing communication latency.

[0071] S4.3: All-to-All Intra-MoE Accelerated Execution (FWD Stage, MoE Model): When Phase_ID=00 (FWD) and Op_Hint=100 (All-to-All) (applicable to the Expert routing communication phase of the Hybrid Expert Model MoE), CFU is configured to perform All-to-All intranet acceleration: Each node will organize the routed Token feature vector into fragments according to the destination Expert node and send them to CFU; CFU performs non-protocol routing operations (CFU_Op_Type=All-to-All), forwarding each fragment directly (Output_Route=PT) to the destination Expert node without performing any protocol calculations; After receiving the corresponding fragment, the target Expert node HCA writes it to the Expert input buffer in the GPU memory for use in MoE forward computation.

[0072] S5: Post-execution status feedback and adaptive update of the strategy table.

[0073] S5.1: CFU Statistical Data Collection: After each aggregation operation is completed, the CFU execution engine reports the following statistics to the AN management processor: actual execution operator type, precision mode, reduction time (nanoseconds), number of participating nodes, and data volume (number of elements). The AN management processor writes the above statistics into the Ring Statistics Buffer (RSB) on the switch side.

[0074] S5.2: Visualization and Strategy Optimization via UFM: NVIDIA Unified Fiber Manager (UFM) periodically polls the statistical data in the RSB of each switch through the InfiniBand subnet management interface. The UFM visualization interface displays the CFU utilization and reduction throughput of each operator type and precision mode under different training stages in the form of a heatmap. Based on the visualization results, the entry configuration in the POMT of each switch can be adjusted, or the optimized policy table can be distributed to all switch nodes in the network through the POMT batch update interface provided by UFM.

[0075] It should be noted that in S2 (S2.2, S2.3), HCA sends PD to the switch-side SMA in advance through the MAD out-of-band channel. On the other hand, the PGM of the HCA firmware layer continuously monitors the GPU computing progress counter. In other words, the computing progress gating (S2.3) does not have to wait until the atomic reconfiguration of the CFU register group (S3.3) is completed before it starts to execute. Instead, it can start monitoring the GPU progress synchronously after PD enters the HCA side sending process.

[0076] However, judging from the formal submission conditions of intra-network aggregate communication execution (such as Reduce-Scatter aggregation operation) on the data plane, this aggregate communication request must simultaneously meet two preconditions: First, the switch side has received the PD, and the AN management processor has completed the stage-operator mapping strategy table lookup, writing the corresponding CFU operator type, precision mode and output routing strategy into the CFU control register group CCRF, and completing the CFU parameter reconfiguration; Second, the HCA side PGM detects that the current completion progress number of the GPU computing stream is greater than or equal to the target token T_L recorded in the Stream_Progress_Token field of the PD, confirming that the gradient calculation of the Lth layer has been completed.

[0077] Only after both of the above conditions are met will PGM send a "ready" enable signal to the NCCL aggregation communication submission queue COSQ, allowing the Reduce-Scatter aggregation group activation request to be formally submitted to the switch through the SHARP RC data plane transmission channel.

[0078] Example: like Figure 5 As shown: Backpropagation phase of the Wanka LLaMA-3 training cluster: In an LLaMA-3 70B model training cluster consisting of 1024 GPU nodes (8 A100 GPUs per node), the FSDP full-shard data parallel strategy is adopted. The cluster has built a two-level Fat-Tree InfiniBand network through NVIDIA Quantum-X800 switches and the above system has been deployed. In each training iteration, the backpropagation phase needs to perform Reduce-Scatter operations on a total of about 140GB of gradients (FP32), and distribute the gradient shards to 8192 parameter responsible nodes.

[0079] S1: Gradient hook triggering and PD construction: When the training iteration enters the backpropagation of the Lth layer, after the PyTorch automatic differentiation engine completes the gradient calculation for this layer, it triggers the post_accumulate_grad_hook registered by PSP. PSP captures this event, sets Phase_ID to 01 (BWD), Op_Hint to 001 (Reduce-Scatter), Prec_Hint to 000 (FP32, preserving gradient precision), Parallel_Mode to 0001 (FSDP), RS_AG_Chain to 0 (only Reduce-Scatter is executed this time, All-Gather will be triggered separately in the optimizer step), and Overlap_Gate to 1 (enabling progress gating). At the same time, PSP inserts a progress marker into the CUDA Stream and obtains the current Stream_Progress_Token value as T_L (representing the token number for the completion of gradient calculation of the Lth layer). PSP pushes the assembled PD into PDSQ.

[0080] S2: PD Encoding and MAD Packaging: The HCA firmware's PD encoding module immediately reads the PD from the PDSQ, encapsulates it into a MAD datagram with the destination address corresponding to the Quantum-X800 switch's SMA, and sends it out through the InfiniBand management port.

[0081] S3: SMA accept and query strategy: The PD arrives at the switch via an out-of-band channel. The switch's SMA receives the PD's MAD datagram before the SHARP data plane RC transmission is established (approximately 1-2 microseconds in advance), and writes the PD content into the PDRB of the AN management processor. The AN management processor reads the PD and constructs a 12-bit index with Phase_ID=01, Op_Hint=001, Prec_Hint=000, and Parallel_Mode=0001. It then queries the POMT to obtain the CFU parameter vector: CFU_Op_Type=Reduce-Scatter, CFU_Prec_Mode=FP32, Output_Route=SR (fragmented routing), and Enable_RS_AG_Pipeline=0.

[0082] S4: CFU register group atomic reconfiguration: The AN management processor detects that the CFU pipeline is currently in an idle gap after the previous batch All-Gather is completed, and immediately writes the above parameter vector to CCRF with an atomic write operation to complete the CFU parameter reconfiguration in the backpropagation phase of this iteration. The switching time is about 10 nanoseconds.

[0083] S5: Computation flow progress gating and data plane submission release: After the PD is pushed into the PD transmission queue PDSQ, the progress gating module PGM of the HCA firmware layer starts to continuously poll the flow progress counter reported by the GPU through the GPUDirect Async interface. At the same time, HCA encapsulates the PD into an out-of-band MAD control message and sends it to the switch-side SMA in advance. The switch AN management processor completes the POMT query and atomic reconfiguration of the CFU control register group CCRF.

[0084] When PGM detects that the current completion progress number of the GPU computing stream is greater than or equal to the target progress number T_L recorded in the Stream_Progress_Token field of the PD, and the corresponding PD has completed the CFU execution parameter configuration on the switch side, PGM sends a "ready" enable signal to the NCCL set communication submission queue COSQ, allowing the current Reduce-Scatter aggregation group activation request to be formally submitted to the switch through the SHARP RC data plane transmission channel; if the Overlap_Gate bit is 0, the GPU computing stream progress gating is skipped, and the submission is released only based on the validity of the PD and the CFU configuration status.

[0085] In this embodiment, S4: CFU register group atomic reconfiguration is written first, and S5 computation flow progress gating is written later. This is described according to the logical completion order before a Reduce-Scatter data plane communication is finally released.

[0086] S6: Intra-network Reduce-Scatter protocol acceleration and distribution: Each compute node's HCA sends its gradient fragments to the AN nodes of the Quantum-X800 aggregation tree via SHARP RC connections according to the fragment index. The CFUs of each level of the AN nodes perform element-wise addition reduction on the corresponding fragments from all child nodes at FP32 precision. The reduced fragments are directly routed to the HCA of the corresponding parameter responsible node along the downlink of the aggregation tree via the Output_Route=SR strategy and written into the gradient receiving buffer in the GPU memory. The entire Reduce-Scatter operation is completed within the network, and the GPUs of each parameter responsible node do not need to perform any additional ring communication steps.

[0087] S7: Statistical Reporting: The CFU reports the execution statistics of this Reduce-Scatter to the AN management processor, including: operator type is Reduce-Scatter, precision mode is FP32, number of participating nodes is 8192, the fragmentation reduction traffic corresponding to approximately 140GB of gradient data processed, and the execution time of this reduction within the network. In the two-level Fat-Tree, XDR 800G InfiniBand network environment of this embodiment, the effective execution time of this Reduce-Scatter on the CFU side can be recorded as approximately 1.8 × 10^8 ns, or approximately 0.18 s.

[0088] The values ​​in this embodiment are exemplary statistical results under specific cluster size, network topology, and link speed, and do not constitute a limitation on the scope of protection of this invention. Under other node sizes, link speeds, aggregation tree depths, or data fragment sizes, the time taken can be actually obtained by the hardware counter of the CFU execution engine and written into the ring statistical buffer RSB on the switch side for UFM to collect and analyze periodically.

[0089] The above embodiments are only for illustrating the technical concept and features of the present invention, and are intended to enable those skilled in the art to understand the content of the present invention and implement it accordingly. They should not be construed as limiting the scope of protection of the present invention. All equivalent changes or modifications made in accordance with the spirit and essence of the present invention should be covered within the scope of protection of the present invention.

Claims

1. A dynamic orchestration system for ensemble communication operators within a SHARP network during the training phase, characterized in that: It includes: Training framework domain: Running on the host side of the GPU computing node, it is responsible for sensing and extracting the current stage state of the training framework, generating standardized training stage descriptors, and transmitting stage semantics to the switch side through the host channel adapter. Host Channel Adapter Domain: Running at the network interface card firmware layer, it is responsible for receiving stage descriptors from the training framework domain. Before the aggregated communication operation is submitted to the SHARP aggregation group, it forwards the stage descriptors to the subnet management agent of the target switch through the out-of-band control channel of the management datagram. It is also responsible for listening to the GPU computing flow progress token. Switch CFU Dynamic Orchestration Domain: Running in the aggregation node management processor inside the switch, it is responsible for receiving stage descriptors from the host channel adapter domain, reconfiguring CFU execution parameters at runtime according to the stage-operator mapping strategy table, and completing parameter switching during the idle gap of the CFU pipeline after the current aggregation operation is completed.

2. The dynamic orchestration system for training-phase perceptual SHARP network intra-network ensemble communication operators according to claim 1, characterized in that: The training framework domain includes: Training framework: Used to perform AI training tasks, and perceives the stage state through gradient hooks and / or interception of the NCCL collection communication library; Stage-aware probe module: Connected to the training framework, it receives raw information from the training framework and performs feature extraction and standardization of training stage descriptors.

3. The dynamic orchestration system for training-phase perceptual SHARP network intra-network ensemble communication operators according to claim 1, characterized in that: The host channel adapter domain includes: PD Encoding and MAD Forwarding Module: Used to receive stage descriptors, encode them into InfiniBand management datagram attributes, and forward them to the subnet management agent of the target switch in advance, through the out-of-band control channel of the high-priority queue, before the submission of the aggregate communication operation. GPU compute stream progress gating module: Used to listen to the progress tokens of the GPU compute stream and generate gating ready signals based on the progress to control the timing of enqueuing of collection communication operations.

4. The dynamic orchestration system for training-phase perceptual SHARP network intra-network ensemble communication operators according to claim 1, characterized in that: The aforementioned switch CFU dynamic orchestration domain includes: PD Reception and Policy Query Module: Used to receive and parse phase descriptors and validity verifications through the subnet management agent, and query the corresponding CFU parameter vector based on the phase-operator mapping policy table; CFU dynamic parameter reconfigurator: connected to the PD receiving and policy query module, used to detect the CFU pipeline idle interval, and write the queried CFU parameter vector into the CFU register group through atomic write operation to dynamically reconfigure the CFU execution parameters; The aggregated function unit execution engine is configured by the CFU dynamic parameter reconfigurator and is used to perform intranet reduction / aggregation calculations on the input aggregated data and output the results according to the configured route.

5. The dynamic orchestration system for training-phase perceptual SHARP network ensemble communication operators, as described in claim 4, is characterized in that: The parameter vector of the stage-operator mapping strategy table includes operator type, precision mode, output route, and chain enable.

6. The dynamic orchestration system for training-phase perceptual SHARP network intra-network ensemble communication operators according to claim 5, characterized in that: The operator types include All-Reduce, Reduce-Scatter, All-Gather, and Reduce-Scatter+All-Gather chained operations; the precision modes include FP32, BF16, FP16, and FP8; and the output routes include broadcast routes, fragmented routes, and direct routes.

7. A method for dynamically arranging aggregate communication operators within a SHARP network during the training phase, characterized in that: The method includes the following steps: S1: Training framework stage awareness and training stage descriptor construction: Aware of the current stage of the training framework, forming a set of original fields for the training stage descriptor; Collect GPU computation flow progress tokens and fill them into the corresponding fields of the stage descriptor; The assembly field is a 32-byte stage descriptor, which is pushed into the send queue on the host channel adapter side. S2: Stage descriptor encoding, transmission, and gating on the host channel adapter side: Read the stage descriptor from the send queue, encode and encapsulate it into a management datagram and send it to the switch subnet management agent, and write it to the stage descriptor receive buffer; poll the GPU progress, and when the progress reaches the target recorded in the stage descriptor, allow the set operation to commit. S3: Switch-side stage descriptor reception, policy query, and CFU parameter reconfiguration: Read the stage descriptor in the receive buffer, check its validity and the aggregation group it belongs to; use the fields in the stage descriptor as indexes to query the stage-operator mapping policy table, detect CFU pipeline idle gaps, and write the queried parameters into the CFU control register group. S4: Intra-network aggregated communication execution after CFU dynamic orchestration: Each node sends data fragments to the CFU, the CFU performs corresponding calculations and routing operations, and sends the results to the corresponding nodes. S5: Post-execution status feedback and strategy table update: CFU reports execution statistics and writes them to the statistics buffer; it uniformly polls and visualizes the statistical data, and adjusts and issues the updated strategy table accordingly.

8. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: When perceiving the current stage of the training framework: relevant fields are extracted through gradient hook interception and / or NCCL set communication group interception to form the original field set of the stage descriptor to be encoded.

9. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: When acquiring GPU compute flow progress tokens: Before enqueuing NCCL collection communication operations, insert a progress token into the current GPU's command queue and fill the stage descriptor with the sequence number of the token as the progress token.

10. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: The phase descriptor read in S2 is the head phase descriptor of the transmission queue with a higher priority than the normal data transmission queue, and it is encoded as the attribute data field of the InfiniBand management datagram, which is a no-reliable management datagram.

11. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: When checking the validity of the phase descriptor and its associated aggregation group: check the Valid flag in the phase descriptor and verify whether the SHARP aggregation group identifier recorded therein belongs to the aggregation tree node range under the responsibility of this switch.

12. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: In S3: When the CFU execution engine is detected to be in an idle gap where the previous aggregation operation has been completed and the next aggregation operation data packet has not yet arrived, the parameter vector obtained from the query is written to the CFU control register group by an atomic write operation to complete the parameter reconfiguration.

13. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 12, is characterized in that: If the operator type and precision mode indicated by the stage descriptor are the same as the current configuration in the CFU control register group, the reconfiguration step is skipped.

14. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: When CFU is configured to perform Reduce-Scatter intranet acceleration: CFU performs element-wise accumulation reduction on the gradient shards sent by each computing node, and sends the reduction results to the corresponding parameter responsible node according to the sharding routing strategy.

15. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: When CFU is configured to perform chained intranet pipeline acceleration of Reduce-Scatter and All-Gather: CFU executes Reduce-Scatter reduction and All-Gather broadcast in sequence. The output of Reduce-Scatter is used as the input of All-Gather. It is executed in a continuous stream mode within the CFU pipeline, and intermediate results are not written back to GPU memory.

16. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: When the CFU is configured to perform All-to-All intra-network acceleration: each node sends the routed Token feature vector to the CFU in fragments according to the destination Expert node. The CFU performs non-standard routing forwarding operations, forwarding each fragment directly to the destination Expert node.

17. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: The statistics in S5 include: the type of operator actually executed, the precision mode, the reduction time, the number of participating nodes, and the amount of data.

18. The method for dynamic orchestration of ensemble communication operators within a perceptual SHARP network during the training phase, as described in claim 7, is characterized in that: In S5: The statistical data of each switch is polled and visualized through the InfiniBand subnet management interface. The configuration of the phase-operator mapping strategy table is adjusted based on the visualization results, and then distributed to all switches in the network through the batch update interface.

Images