Graphics processor memory access feature data collection method and apparatus

Abstract

This invention provides a method and apparatus for acquiring memory access characteristic data of a graphics processing unit (GPU), relating to the field of computer technology. The method includes: intercepting call parameters of application programming interfaces (APIs) in the GPU runtime library; constructing a memory region tree structure based on the call parameters; inserting a preset function into the execution instruction stream of the GPU kernel to acquire instruction-level execution data; and generating GPU memory access characteristic data based on the memory region tree structure and the instruction-level execution data. The GPU memory access characteristic data acquisition method and apparatus provided by this invention correlate macroscopic memory and communication semantics with microscopic instruction-level execution data to generate global, fine-grained memory access characteristic data, improving the comprehensiveness and accuracy of acquiring GPU memory access characteristic data in cloud-edge-device architectures.

Description

Technical Field

[0001] This invention relates to the field of computer technology, and in particular to a method and apparatus for acquiring memory access characteristic data of a graphics processor. Background Technology

[0002] Currently, the cloud-edge-device architecture (where cloud refers to cloud-based central computing service facilities, edge refers to edge agile access clusters, and device refers to terminal access devices) has become a common topology for performing tasks such as large-scale model training and inference, and scientific computing. In this cloud-edge-device architecture, accurate acquisition of data such as memory access, communication, and scheduling status of graphics processing units (GPUs) in large-scale cluster environments is crucial for optimizing task scheduling, network topology, and resource allocation.

[0003] In existing technologies, GPU behavioral data acquisition methods primarily involve inserting Extended Berkeley Packet Filter (eBPF) probes into user-mode runtime library interfaces and kernel general subsystem functions to collect key behavioral data throughout the GPU lifecycle, generating events such as task submission and memory management. However, this method suffers from the following drawbacks: it only involves data acquisition at the Application Programming Interface (API) call level, lacking analysis at the GPU instruction execution level. Furthermore, it only focuses on local memory management in the memory domain and optimizes the scheduling control of a single GPU in the result analysis, leading to incomplete and low-precision data. Summary of the Invention

[0004] This invention provides a method and apparatus for acquiring memory access feature data of a graphics processor, in order to solve the technical problem that the acquisition of memory access feature data of GPUs in cloud-edge architecture is incomplete and has low accuracy in the prior art.

[0005] This invention provides a method for acquiring memory access characteristic data of a graphics processor, comprising the following steps: Intercept the call parameters of the application programming interface in the graphics processor runtime library; Based on the calling parameters, construct a memory interval tree structure; Preset functions are inserted into the instruction stream of the graphics processor kernel to collect instruction-level execution data; Based on the memory interval tree structure and the instruction-level execution data, memory access characteristic data of the graphics processor is generated.

[0006] According to a method for acquiring memory access characteristic data of a graphics processor provided by the present invention, the instruction-level execution data includes at least one of the following: The memory address of each memory access instruction in the execution instruction stream; Timing information of each memory access instruction in the execution instruction stream; The execution context information of each memory access instruction in the execution instruction stream.

[0007] According to a method for acquiring memory access characteristic data of a graphics processing unit (GPU) provided by the present invention, generating memory access characteristic data of the GPU based on the memory interval tree structure and the instruction-level execution data includes: Based on the memory access address, a query and match are performed in the memory interval tree structure to obtain the memory semantic information associated with each memory access instruction; Based on the memory semantic information, a memory access record corresponding to each memory access instruction is generated; Based on the timestamps in the timing information, the memory access records are scanned sequentially to construct the memory access dependency relationship between the memory access operations corresponding to each memory access instruction; Based on the memory access dependency relationship, the memory access feature data is generated.

[0008] According to the present invention, a method for acquiring memory access characteristic data of a graphics processor is provided, wherein the calling parameters include at least one of the following: Memory allocation functions; Memory release function; Host memory fixed functions; Data communication functions.

[0009] According to a method for acquiring memory access characteristic data of a graphics processor provided by the present invention, the data communication function includes at least one of the following: Memory copy function; Point-to-point communication functions; Memory registration function for remote direct memory access; Communication functions of the message passing interface.

[0010] According to a method for acquiring memory access characteristic data of a graphics processor provided by the present invention, the memory interval tree structure is used to describe non-overlapping virtual memory intervals, and the data in the memory interval tree structure includes at least one of the following: The starting endpoint of the virtual memory region; The memory type of the virtual memory region; The pointer to the interval information record structure of the virtual memory interval.

[0011] The present invention also provides a memory access feature data acquisition device for a graphics processor, comprising the following modules: The interception module is used to intercept the call parameters of the application programming interface in the graphics processor runtime library; The construction module is used to construct a memory range tree structure based on the calling parameters; The acquisition module is used to insert preset functions into the execution instruction stream of the graphics processor kernel and acquire instruction-level execution data; The generation module is used to generate memory access characteristic data of the graphics processor based on the memory range tree structure and the instruction-level execution data.

[0012] The present invention also provides an electronic device, including a memory, a processor, and a computer program stored in the memory and executable on the processor, wherein the processor executes the computer program to implement a memory access feature data acquisition method for any of the graphics processors described above.

[0013] The present invention also provides a non-transitory computer-readable storage medium having a computer program stored thereon, which, when executed by a processor, implements a memory access feature data acquisition method for a graphics processor as described above.

[0014] The present invention also provides a computer program product, including a computer program that, when executed by a processor, implements a memory access feature data acquisition method for a graphics processor as described above.

[0015] This invention provides a method and apparatus for acquiring memory access characteristic data of a graphics processing unit (GPU). By intercepting the call parameters of the application programming interface (API) in the GPU runtime library, it captures the original semantics of memory operations and communication requests initiated by the GPU program, providing a data foundation for subsequently constructing a global, unified memory view. Based on the call parameters, a memory interval tree structure is constructed, thereby unifying scattered and independent memory allocation and communication operations into a continuous address interval, achieving an integrated description of physical memory, virtualized memory, and communication buffers in complex cloud-edge-device environments. Preset functions are inserted into the GPU kernel's instruction stream to collect instruction-level execution data, overcoming the limitations of monitoring only at the software API level. This directly collects instruction-level execution data for each memory access instruction executed by the GPU hardware stream multiprocessor, achieving high-precision observation of the actual behavior of the computing core. Based on the memory interval tree structure and instruction-level execution data, GPU memory access characteristic data is generated, thus associating macroscopic memory and communication semantics with microscopic instruction-level execution data to generate global, fine-grained memory access characteristic data, improving the comprehensiveness and accuracy of acquiring GPU memory access characteristic data in cloud-edge-device architectures. Attached Figure Description

[0016] To more clearly illustrate the technical solutions in this invention or the prior art, the drawings used in the description of the embodiments or the prior art will be briefly introduced below. Obviously, the drawings described below are some embodiments of this invention. For those skilled in the art, other drawings can be obtained from these drawings without creative effort.

[0017] Figure 1 This is a flowchart illustrating the memory access feature data acquisition method for a graphics processor provided by the present invention.

[0018] Figure 2 This is a schematic diagram of the structure of the memory access feature data acquisition device for the graphics processor provided by the present invention.

[0019] Figure 3 This is a schematic diagram of the structure of the electronic device provided by the present invention. Detailed Implementation

[0020] To make the objectives, technical solutions, and advantages of this invention clearer, the technical solutions of this invention will be clearly and completely described below with reference to the accompanying drawings. Obviously, the described embodiments are only some, not all, of the embodiments of this invention. All other embodiments obtained by those skilled in the art based on the embodiments of this invention without creative effort are within the scope of protection of this invention.

[0021] It should be noted that, in this document, the terms "comprising," "including," or any other variations thereof are intended to cover non-exclusive inclusion, such that a process, method, article, or apparatus that comprises a list of elements includes not only those elements but also other elements not expressly listed, or elements inherent to such a process, method, article, or apparatus. Without further limitations, an element defined by the phrase "comprising one..." does not exclude the presence of other identical elements in the process, method, article, or apparatus that includes that element. Furthermore, it should be noted that the scope of the methods and apparatuses in the embodiments of this application is not limited to performing functions in the order shown or discussed, but may also include performing functions substantially simultaneously or in the reverse order, depending on the functions involved. For example, the described methods may be performed in a different order than described, and various steps may be added, omitted, or combined. Additionally, features described with reference to certain examples may be combined in other examples.

[0022] In this application's embodiments, "determine B based on A" means that factor A must be considered when determining B. It is not limited to "B can be determined based solely on A," but should also include: "determine B based on A and C," "determine B based on A, C, and E," "determine C based on A, and further determine B based on C," etc. Additionally, it can include using A as a condition for determining B, for example, "when A meets the first condition, determine B using the first method"; another example, "when A meets the second condition, determine B," etc.; another example, "when A meets the third condition, determine B based on the first parameter," etc. Of course, it can also be a condition where A is a factor in determining B, for example, "when A meets the first condition, determine C using the first method, and further determine B based on C," etc.

[0023] It should also be noted that the terms "target," "first," and "second" in this invention are used to distinguish similar objects, not to describe a specific order or sequence. It should be understood that such terms can be used interchangeably where appropriate so that embodiments of this application can be implemented in orders other than those illustrated or described herein, and the objects distinguished by "first" and "second" are generally of the same class, without limiting the number of objects; for example, the first object can be one or more.

[0024] In this invention, the term "multiple" refers to two or more, and other quantifiers are similar.

[0025] The following is combined Figures 1 to 3 The present invention describes a method and apparatus for acquiring memory access characteristic data of a graphics processor.

[0026] Figure 1 This is a flowchart illustrating the memory access feature data acquisition method for a graphics processor provided by the present invention, as shown below. Figure 1 As shown, the method includes the following steps: Step 101: Intercept the call parameters of the application programming interface in the graphics processor runtime library; Specifically, by intercepting (or hijacking) all critical memory and communication operation instructions and their parameters initiated by the application and directed towards the GPU, the call parameters of the original memory operation requests issued by the application can be obtained. The interception methods can include dynamic link library-based hijacking methods, such as Linux LD_PRELOAD and Windows MinHook; or kernel-mode probe-based hijacking methods, such as eBPF probes.

[0027] When implementing function interception in a dynamic linking environment, firstly, a proxy function prototype with a signature identical to the target intercepted function is constructed to ensure correct linker binding. Secondly, within the proxy function, a global function pointer is defined, and on the first call, the address of the original function is obtained and saved using the `dlsym` function in conjunction with the `RTLD_NEXT` handle. Next, custom pre-processing and post-processing code is inserted into the proxy function body to implement specific interception logic. Finally, the code containing this proxy function is compiled into a position-independent shared object (dynamic link library), and the library is forcibly preloaded using the operating system's dynamic linking environment variable (such as Linux's `LD_PRELOAD`), thus giving the function symbols within it higher priority than the original symbols in the application and its dependent libraries, achieving function interception and replacement.

[0028] Furthermore, the calling parameters include at least one of the following: Memory allocation functions; Memory release function; Host memory fixed functions; Data communication functions.

[0029] Furthermore, the data communication function includes at least one of the following: Memory copy function; Point-to-point communication functions; Memory registration function for remote direct memory access; Communication functions of the message passing interface.

[0030] Specifically, the intercepted call parameters may include one or more of the following: memory allocation functions, memory release functions, host memory anchoring functions, or data communication functions. Among these, data communication functions include one or more of the following: memory copy functions, point-to-point communication functions, remote direct memory access memory registration functions, or message passing interface communication functions.

[0031] For example, in one embodiment, a dynamic link library hijacking method based on LD_PRELOAD is used to intercept the memory copy function (cudaMemcpy) in the NVIDIA CUDA runtime library. First, a proxy function with a signature exactly identical to the original cudaMemcpy function is written. When the application calls cudaMemcpy(dst, src, size, kind), the operating system's dynamic linker will preferentially link to this proxy function. Inside the proxy function, the address of the original function is obtained and saved using dlsym(RTLD_NEXT, "cudaMalloc"). Then, before and after calling the original function, pre-processing code (recording call parameters, such as the requested memory size) and post-processing code (recording the returned memory pointer ptr) are executed respectively. Finally, by setting LD_PRELOAD=. / libmem_hook.so, this proxy library is preferentially loaded during application runtime, thereby intercepting all calls to cudaMalloc.

[0032] This invention achieves non-intrusive and transparent monitoring of application-layer GPU operations by intercepting all memory operation intentions and key parameters (such as address and size) initiated through standard APIs. It can completely capture the intentions and context of all memory allocation, release, and data transmission, and is applicable to various complex scenarios in cloud-edge-device architectures such as single-card acquisition, single-machine multi-card acquisition, and multi-machine multi-card acquisition, meeting the needs of complex computing architecture evolution.

[0033] Step 102: Based on the calling parameters, construct a memory region tree structure; Specifically, based on the various memory address requests in the call parameters, non-overlapping virtual memory regions are obtained, and then a red-black tree is used to maintain the right endpoint address of the region to construct a memory region tree structure.

[0034] Furthermore, the memory interval tree structure is used to describe non-overlapping virtual memory intervals, and the data in the memory interval tree structure includes at least one of the following: The starting endpoint of the virtual memory region; The memory type of the virtual memory region; The pointer to the interval information record structure of the virtual memory interval.

[0035] Specifically, the memory interval tree structure includes the following fields: 1) The starting and ending points of the virtual memory region; 2) The memory type of the virtual memory region. This field is a binary marker and includes, but is not limited to, the following fields: MEM_DEVICE_LOCAL: Memory allocated locally on the GPU (0x1); MEM_HOST_LOCAL: NUMA memory allocated by the CPU (0x2); MEM_CPY_SRC: The memory source (0x100) transferred from the Device to the Host. MEM_CPY_DST: Memory destination (0x200) transferred from Device to Host; MEM_P2P_SRC: Device memory source for mutual transfer between multiple GPUs (0x400); MEM_P2P_DST: Device memory target (0x800) for inter-GPU data transfer. MEM_RDMA_ACCESS: Memory segment mapped to RDMA (0x00010000). MEM_GDS_ACCESS: Memory segment mapped to GDS (0x00020000); MEM_MPI_ACCESS: Memory segment mapped to MPI (0x00040000); 3) A pointer to the record structure of the virtual memory region, used to point to the record structure of the current region; Each time a range attribute is updated, a new range record structure is generated. It should be noted that when a part of a large range is updated to a remote communication domain, the original range needs to be split into multiple ranges to ensure that the memory type semantics of the ranges are unambiguous, and the relevant records in the red-black tree are updated at the same time.

[0036] All interval information is maintained using a variable-length array, and each interval information includes, but is not limited to, the following fields: 1) A range allocation transaction identifier (ID) is used to mark which API call generated the current range; 2) The transaction ID before the range change, used to identify which transaction changed the current range; 3) Allocate GPU timestamps in intervals for timing alignment with the acquisition results of the NVIDIA Binary Instrumentation Tool (NVBit); 4) The GPU timestamps released in the interval are used for time-series alignment with the acquisition results of NVBit; 5) A hash table of visited page information, used to map the Non-Uniform Memory Access (NUMA) node number to which the current page is mapped; 6) Remote communication status description pointer, used to record the remote communication status; 7) Memory access dependency table, used to record the memory access status of this data.

[0037] For example, in one embodiment, after intercepting the cudaMemcpy function call and recording its call parameters, the system constructs and maintains a memory range tree structure organized in a red-black tree (i.e., an existing self-balancing binary search tree data structure): the source memory range [src, src+size) is determined by the source address src and size size; the node of this range is searched and updated in the red-black tree, and MEM_CPY_SRC (indicating that this range is being used as the source for memory copying) is added to its binary tag field. Similarly, the MEM_CPY_DST tag is added to the target address range [dst, dst+size). If this copy operation only involves a part of an existing large range, a red-black tree node splitting operation is triggered, thereby ensuring that the address range represented by each tree node has a unique and clear semantic. In each node of the red-black tree (representing a memory range), the following key information is also stored: 1) Starting endpoint: the virtual address at the beginning of the memory range (e.g., 0x1000); 2) Memory type: a binary bit tag field used to encode various semantic attributes of the range. For example, an interval may have both MEM_DEVICE_LOCAL (device local) and MEM_CPY_SRC (copy source) flags; 3) Interval information record structure pointer: used to point to a detailed dynamic record structure, which contains historical and related information such as transaction ID, timestamp, parent transaction pointer, and memory access dependency table.

[0038] This invention manages all non-overlapping virtual memory regions by constructing an efficient memory region tree structure and assigning rich semantics to each region using binary tags. This achieves unified view management of multi-type, cross-domain memory, ensuring that region insertion, deletion, search, and splitting operations are completed in logarithmic time complexity, reducing performance overhead and improving processing efficiency. Simultaneously, the region information record structure pointer decouples static region definitions from dynamic change history, recording the complete lifecycle of each region (including which transaction created it, which transaction modified it, and which memory accesses it is related to), providing a foundation for subsequently constructing accurate memory access dependencies.

[0039] Step 103: Insert a preset function into the execution instruction stream of the graphics processor kernel and collect instruction-level execution data; Specifically, by dynamically inserting probe code (i.e., preset functions) into the instruction stream of the GPU kernel, fine-grained execution information of each memory access instruction actually executed by the Streaming Multiprocessor (SM) is directly captured as instruction-level execution data.

[0040] Furthermore, the instruction-level execution data includes at least one of the following: The memory address of each memory access instruction in the execution instruction stream; Timing information of each memory access instruction in the execution instruction stream; The execution context information of each memory access instruction in the execution instruction stream.

[0041] Specifically, instruction-level execution data includes one or more of the following: the memory address, timing information, and execution context information of each memory access instruction in the execution instruction stream of the graphics processor core, including but not limited to the following: 1) The target acquisition address for each thread, which is of type uint64

[32] and is used to record each memory address information; 2) Active thread ID, used to filter the actual effective address during the execution of each memory access instruction; 3) Data read bit width, which is the width of data read and write in the stream processor assembly SASS instruction set (i.e., the underlying machine instruction set of NVIDIA GPU, including but not limited to load LD, store ST, global load LDG, global store STG, local load LDL, local store STL, atomic operation ATOM, global atomic operation ATOMG, reduction atomic double word operation RAD, etc. stream processor assembly SASS instructions). 4) Thread bundle warp local ID, that is, the ID of the thread bundle warp in a cooperative thread array (CTA); 5) CTA ID, which is the ID of the CTA in the entire SM. This field is used together with the thread bundle warp local ID to identify a unique index in a kernel execution; 6) SM ID, i.e., the SM issued by the current command; 7) GPU time, which is the SM clock of the current instruction.

[0042] For example, in one embodiment, by hijacking the kernel startup function cuLaunchKernel, a binary instrumentation tool (such as NVIDIA's NVBit) is used to analyze the stream processor assembly SASS instructions before the target GPU kernel function is executed, and a custom preset function is injected before each memory access instruction (such as the global load LDG instruction). The default function collects and encapsulates the following information: 1) Memory address: The memory address to be read obtained from the actual executing thread through the thread shuffling function _shfl_sync, for example, 0xB0000000; 2) Timing information: The clock count obtained by reading the GPU's %globaltimer register, and the accurate timestamp aligned with the CPU time by calculating tc0 + (tg - tg0) based on the baseline CPU timestamp (tc0) and baseline GPU timestamp (tg0) recorded when the kernel function starts; 3) Execution context information: Including the local ID of the thread bundle that issued the instruction in its thread block (or CTA), the ID of the thread block in the entire grid, and the ID of the streaming multiprocessor that executed the instruction.

[0043] This invention improves the comprehensiveness and accuracy of collecting memory access characteristic data of GPUs in cloud-edge-device architecture by acquiring the micro-execution trajectory of GPU hardware, including precise memory access addresses, thread context, and high-precision timestamps.

[0044] Step 104: Based on the memory interval tree structure and the instruction-level execution data, generate memory access characteristic data of the graphics processor.

[0045] Specifically, macroscopic, semantic memory interval trees are associated, matched, and comprehensively analyzed with microscopic, precise instruction-level data to generate global memory access characteristic data.

[0046] Furthermore, the step of generating memory access characteristic data of the graphics processor based on the memory interval tree structure and the instruction-level execution data includes: Based on the memory access address, a query and match are performed in the memory interval tree structure to obtain the memory semantic information associated with each memory access instruction; Based on the memory semantic information, a memory access record corresponding to each memory access instruction is generated; Based on the timestamps in the timing information, the memory access records are scanned sequentially to construct the memory access dependency relationship between the memory access operations corresponding to each memory access instruction; Based on the memory access dependency relationship, the memory access feature data is generated.

[0047] Specifically, the first step is to align the timestamps of the instruction-level execution data, which includes: 1) Clock information reading. A static sm_clock array is maintained on the Central Processing Unit (SM) side, representing the time information maintained by each SM. It contains two fields: the GPU time at which each kernel starts execution and the number of GPU cycles, which serve as the calibration benchmark GPU clock and benchmark GPU cycle. 2) Data Synchronization with the CPU. A fixed circular buffer is established, allocated on the host and mapped to the device memory. Data is continuously written to this circular buffer, using the above data structure as the acquisition unit. An atomic pointer (head) and a pointer (tail) are maintained for this circular buffer. When writing or reading data, the head is atomically updated first, then the data is copied, and finally the tail is atomically updated. The CPU is notified when the buffer is full. A separate receiving thread is used on the CPU side, maintaining a shared hint variable. This variable is constantly polled within the thread to perform the corresponding read operations. To ensure that GPU writes are unique per warp, it is always agreed that the smallest active thread performs the actual write operation. The smallest thread ID in the active mask can be obtained using the `__ffs` function.

[0048] Once all instruction-level execution data for a kernel has been collected and timestamp aligned, a query and match is performed in the memory region tree structure based on the memory access address to obtain the memory semantic information associated with each memory access instruction. This involves adding the following fields as remote memory access information: Memory access type: Read (L) / Store (W); Memory access address: The address accessed during memory access; Memory access corresponds to the memory type: Device / Host / Remote; Is this the first time accessing a memory page: True / False; Memory access address dependency: The LDX address corresponds to the API Call ID of the last write operation; The memory access address is dependent on the API Call ID used for the first time after the STX address is written.

[0049] For example, for LDX / STX statements, the accessed memory semantic information can be queried in the memory interval tree structure. For instance, if the lower_bound binary search operation is called at the right endpoint of the interval, the first interval found is the corresponding memory access interval.

[0050] For example, the actual usage of host memory allocated on a Non-Uniform Memory Access (NUMA) node is deferred until an actual memory access occurs. To extract its instruction information, an additional NUMAPageMap needs to be maintained.<PageId,NodeId> This is a page table mapping. Therefore, assuming the memory access location is L, then [L / 32] is the actual page address retrieved. Before collecting the page address, the NUMA node information in the system is first searched to determine whether the corresponding ID is remote or near. During the page address collection process, if L exists in NUMAPageMap, the NodeId is directly returned as the memory access node; if L does not exist in NUMAPageMap, get_mempolicy is called to obtain the corresponding memory access address information and update it in NUMAPageMap.

[0051] Based on the retrieved memory semantic information, generate or update the memory access record corresponding to each memory access instruction, and scan the new memory access records sequentially using the timestamps in the timing information to construct the memory access dependency relationship (such as the memory access dependency table) between the memory access operations corresponding to each memory access instruction.

[0052] For example, if the memory access instruction type is read, the memory address of the instruction is searched in the memory range tree structure. If the memory range tree structure does not contain a range containing the instruction, it is skipped; otherwise, the timestamp of the corresponding range is searched. If the timestamp is later than the execution timestamp of the instruction, the parent transaction ID is searched continuously based on the range information until the first timestamp earlier than the execution timestamp of the instruction is found. If not found, it is skipped. If found, and the range has no communication marker, it is skipped. If there is a communication request, the instruction number is added to the memory access dependency table.

[0053] For example, if the memory access instruction type is write, the memory address of the instruction is searched in the memory range tree structure. If the memory range tree structure does not contain a range containing the instruction, it is skipped; otherwise, the timestamp of the corresponding range is searched. If the timestamp is earlier than the execution timestamp of the instruction, it is skipped. The parent transaction ID is continuously searched based on the range information until a timestamp later than the execution of the instruction and earlier than the execution of the parent transaction is found. If not found, it is skipped. If found, and the range has no communication marker, it is skipped. If there is a communication marker, all instruction numbers are traversed in the memory access dependency table. If there is no write to the same address, it is added. If there is a write to the same address, the write dependency of the corresponding instruction is removed, the original instruction number is removed from the memory access dependency table, and the current instruction number is added.

[0054] The memory access dependency table records memory access feature data with rich semantics, such as: "In Warp 5 on SM2, a load operation on device memory is based on data from a memory copy from host to device (cudaMemcpy) 1.2ms ago, which was initiated by API call ID #123." Based on the above embodiments, this invention establishes memory access dependencies between memory access operations corresponding to each memory access instruction through timing analysis, generating global memory access feature data that integrates memory semantics, instruction trajectories, and data flow dependencies. This clearly reveals the data flow path between the memory plane and the communication plane, achieving "dual-plane" information fusion. As a result, the final generated memory access feature data not only contains static attributes but also dynamic data flow relationships, thereby improving the comprehensiveness and accuracy of memory access feature data collection for GPUs in cloud-edge-device architectures.

[0055] This invention provides a method for acquiring memory access characteristic data of a graphics processing unit (GPU). By intercepting the call parameters of the application programming interface (API) in the GPU runtime library, it captures the original semantics of memory operations and communication requests initiated by the GPU program, providing a data foundation for subsequently constructing a global, unified memory view. Based on the call parameters, a memory interval tree structure is constructed, thereby unifying scattered and independent memory allocation and communication operations into a continuous address interval, achieving an integrated description of physical memory, virtualized memory, and communication buffers in complex cloud-edge-device environments. Preset functions are inserted into the GPU kernel's instruction stream to collect instruction-level execution data, overcoming the limitations of monitoring only at the software API level. This directly collects instruction-level execution data for each memory access instruction executed by the GPU hardware stream multiprocessor, achieving high-precision observation of the actual behavior of the computing core. Based on the memory interval tree structure and instruction-level execution data, GPU memory access characteristic data is generated, thus associating macroscopic memory and communication semantics with microscopic instruction-level execution data to generate global, fine-grained memory access characteristic data, improving the comprehensiveness and accuracy of GPU memory access characteristic data acquisition in cloud-edge-device architectures.

[0056] The memory access feature data acquisition device for a graphics processor provided by the present invention will be described below. The memory access feature data acquisition device for a graphics processor described below can be referred to in correspondence with the memory access feature data acquisition method for a graphics processor described above.

[0057] Figure 2 This is a schematic diagram of the structure of the memory access feature data acquisition device for the graphics processor provided by the present invention, as shown below. Figure 2 As shown. An embodiment of the present invention provides a memory access feature data acquisition device for a graphics processor, comprising an interception module 201, a construction module 202, an acquisition module 203, and a generation module 204, wherein: The interception module 201 is used to intercept the call parameters of the application programming interface in the graphics processor runtime library; the construction module 202 is used to construct a memory range tree structure based on the call parameters; the acquisition module 203 is used to insert a preset function into the execution instruction stream of the graphics processor kernel and acquire instruction-level execution data; the generation module 204 is used to generate memory access characteristic data of the graphics processor based on the memory range tree structure and the instruction-level execution data.

[0058] This invention provides a memory access characteristic data acquisition device for graphics processing units (GPUs). By intercepting the call parameters of application programming interfaces (APIs) in the GPU runtime library, it captures the original semantics of memory operations and communication requests initiated by GPU programs, providing a data foundation for subsequently constructing a global, unified memory view. Based on the call parameters, a memory interval tree structure is constructed, thereby unifying scattered and independent memory allocation and communication operations into a continuous address interval, achieving an integrated description of physical memory, virtualized memory, and communication buffers in complex cloud-edge-device environments. Preset functions are inserted into the GPU kernel's instruction stream to collect instruction-level execution data, thus overcoming the limitations of monitoring only at the software API level and directly collecting instruction-level execution data for each memory access instruction executed by the GPU hardware stream multiprocessor, achieving high-precision observation of the actual behavior of the computing core. Based on the memory interval tree structure and instruction-level execution data, GPU memory access characteristic data is generated, thereby associating macroscopic memory and communication semantics with microscopic instruction-level execution data to generate global, fine-grained memory access characteristic data, improving the comprehensiveness and accuracy of GPU memory access characteristic data acquisition in cloud-edge-device architectures.

[0059] Figure 3 An example is a schematic diagram of the physical structure of an electronic device, such as... Figure 3 As shown, the electronic device may include: a processor 310, a communications interface 320, a memory 330, and a communication bus 340, wherein the processor 310, the communications interface 320, and the memory 330 communicate with each other via the communication bus 340. The processor 310 can call logical instructions in the memory 330 to execute a memory access feature data acquisition method for a graphics processor, the method including: Intercept the call parameters of the application programming interface in the graphics processor runtime library; Based on the calling parameters, construct a memory interval tree structure; Preset functions are inserted into the instruction stream of the graphics processor kernel to collect instruction-level execution data; Based on the memory interval tree structure and the instruction-level execution data, memory access characteristic data of the graphics processor is generated.

[0060] Furthermore, the logical instructions in the aforementioned memory 330 can be implemented as software functional units and, when sold or used as independent products, can be stored in a computer-readable storage medium. Based on this understanding, the technical solution of the present invention, essentially, or the part that contributes to the prior art, or a part of the technical solution, can be embodied in the form of a software product. This computer software product is stored in a storage medium and includes several instructions to cause a computer device (which may be a personal computer, server, or network device, etc.) to execute all or part of the steps of the methods described in the various embodiments of the present invention. The aforementioned storage medium includes various media capable of storing program code, such as USB flash drives, portable hard drives, read-only memory (ROM), random access memory (RAM), magnetic disks, or optical disks.

[0061] On the other hand, the present invention also provides a computer program product, the computer program product comprising a computer program that can be stored on a non-transitory computer-readable storage medium, wherein when the computer program is executed by a processor, the computer is able to execute the graphics processor memory access feature data acquisition method provided by the above methods, the method comprising: Intercept the call parameters of the application programming interface in the graphics processor runtime library; Based on the calling parameters, construct a memory interval tree structure; Preset functions are inserted into the instruction stream of the graphics processor kernel to collect instruction-level execution data; Based on the memory interval tree structure and the instruction-level execution data, memory access characteristic data of the graphics processor is generated.

[0062] In another aspect, the present invention also provides a non-transitory computer-readable storage medium having a computer program stored thereon, which, when executed by a processor, implements a method for acquiring memory access feature data of a graphics processor provided by the methods described above, the method comprising: Intercept the call parameters of the application programming interface in the graphics processor runtime library; Based on the calling parameters, construct a memory interval tree structure; Preset functions are inserted into the instruction stream of the graphics processor kernel to collect instruction-level execution data; Based on the memory interval tree structure and the instruction-level execution data, memory access characteristic data of the graphics processor is generated.

[0063] The device embodiments described above are merely illustrative. The units described as separate components may or may not be physically separate. The components shown as units may or may not be physical units; that is, they may be located in one place or distributed across multiple network units. Some or all of the modules can be selected to achieve the purpose of this embodiment according to actual needs. Those skilled in the art can understand and implement this without any creative effort.

[0064] Through the above description of the embodiments, those skilled in the art can clearly understand that each embodiment can be implemented by means of software plus necessary general-purpose hardware platforms, and of course, it can also be implemented by hardware. Based on this understanding, the above technical solutions, in essence or the part that contributes to the prior art, can be embodied in the form of a software product. This computer software product can be stored in a computer-readable storage medium, such as ROM / RAM, magnetic disk, optical disk, etc., and includes several instructions to cause a computer device (which may be a personal computer, server, or network device, etc.) to execute the methods described in the various embodiments or some parts of the embodiments.

[0065] Finally, it should be noted that the above embodiments are only used to illustrate the technical solutions of the present invention, and not to limit them; although the present invention has been described in detail with reference to the foregoing embodiments, those skilled in the art should understand that modifications can still be made to the technical solutions described in the foregoing embodiments, or equivalent substitutions can be made to some of the technical features; and these modifications or substitutions do not cause the essence of the corresponding technical solutions to deviate from the spirit and scope of the technical solutions of the embodiments of the present invention.

Claims

1. A method for collecting memory access characteristic data of a graphics processor, the method comprising: include: Intercept the call parameters of the application programming interface in the graphics processor runtime library; Based on the calling parameters, construct a memory interval tree structure; Preset functions are inserted into the instruction stream of the graphics processor kernel to collect instruction-level execution data; Based on the memory interval tree structure and the instruction-level execution data, memory access characteristic data of the graphics processor is generated.

2. The method according to claim 1, wherein, The instruction-level execution data includes at least one of the following: The memory address of each memory access instruction in the execution instruction stream; Timing information of each memory access instruction in the execution instruction stream; The execution context information of each memory access instruction in the execution instruction stream.

3. The method according to claim 2, wherein, The generation of memory access characteristic data for the graphics processor based on the memory interval tree structure and the instruction-level execution data includes: Based on the memory access address, a query and match are performed in the memory interval tree structure to obtain the memory semantic information associated with each memory access instruction; Based on the memory semantic information, a memory access record corresponding to each memory access instruction is generated; Based on the timestamps in the timing information, the memory access records are scanned sequentially to construct the memory access dependency relationship between the memory access operations corresponding to each memory access instruction; Based on the memory access dependency relationship, the memory access feature data is generated.

4. The method of claim 1, wherein the method further comprises: The calling parameters include at least one of the following: Memory allocation functions; Memory release function; Host memory fixed functions; Data communication functions.

5. The method for acquiring memory access characteristic data of a graphics processor according to claim 4, characterized in that, The data communication function includes at least one of the following: Memory copy function; Point-to-point communication functions; Memory registration function for remote direct memory access; Communication functions of the message passing interface.

6. The method for acquiring memory access feature data of a graphics processor according to claim 1, characterized in that, The memory interval tree structure is used to describe non-overlapping virtual memory intervals, and the data in the memory interval tree structure includes at least one of the following: The starting endpoint of the virtual memory region; The memory type of the virtual memory region; The pointer to the interval information record structure of the virtual memory interval.

7. A memory access characteristic data acquisition device for a graphics processor, characterized in that, include: The interception module is used to intercept the call parameters of the application programming interface in the graphics processor runtime library; The construction module is used to construct a memory range tree structure based on the calling parameters; The acquisition module is used to insert preset functions into the execution instruction stream of the graphics processor kernel and acquire instruction-level execution data; The generation module is used to generate memory access characteristic data of the graphics processor based on the memory range tree structure and the instruction-level execution data.

8. An electronic device comprising a memory, a processor, and a computer program stored in the memory and running on the processor, characterized in that, When the processor executes the computer program, it implements the memory access feature data acquisition method of the graphics processor as described in any one of claims 1 to 6.

9. A non-transitory computer-readable storage medium having a computer program stored thereon, characterized in that, When the computer program is executed by the processor, it implements the memory access feature data acquisition method of the graphics processor as described in any one of claims 1 to 6.

10. A computer program product, comprising a computer program, characterized in that, When the computer program is executed by the processor, it implements the memory access feature data acquisition method of the graphics processor as described in any one of claims 1 to 6.

Images