Implementation method and system of ONNX deep learning operator on OpenCL equipment and medium

Abstract

The invention discloses an implementation method and system of an ONNX deep learning operator on OpenCL equipment and a medium, and the method comprises the steps: configuring an OpenCL ONNX operator, executing the OpenCL ONNX operator on the OpenCL equipment according to the received parameters required by a target operator, and executing the operator operation on a GPU in the execution process, and after the operator operation is executed, judging whether the operator operation is transmitted to the CPU end or kept in the GPU for next operator operation according to an execution result. According to the method, efficient and general calculation support of an ONNX operator on OpenCL equipment can be realized, the dependence of a deep learning framework on single hardware is broken through, the calculation overhead during multi-dimensional tensor calculation is reduced, and the applicability of calculation of different shapes of tensors can be improved.

Description

Technical Field

[0001] The present invention relates to the field of artificial intelligence technology, and particularly to a method, system and medium for implementing ONNX deep learning operators on OpenCL devices. Background Art

[0002] With the rapid development of artificial intelligence (AI) technology, deep learning has become a key technology driving the development of intelligent applications, especially in the fields of image processing, speech recognition, and natural language processing. Deep learning frameworks usually rely on high-performance computing devices to accelerate model training and inference to meet the growing data processing requirements. To achieve cross-platform and cross-device compatibility and efficient computing, different hardware manufacturers have developed operator libraries that support specific standards (such as CUDA, ROCM, SYCL, OpenCL). However, current deep learning frameworks mainly support CUDA or specific manufacturers' computing acceleration libraries, which limits the general support for heterogeneous computing devices (such as various GPUs, FPGAs, NPUs, ASICs, etc. computing cards). When running on proprietary hardware, deep learning frameworks need to perform specific data structure conversions on input and output tensors, making it difficult to adapt to general computing devices and resulting in poor scalability. For example, a method, device, equipment and medium for accessing a chip operator library into a deep learning framework disclosed in Chinese patent document with application number CN202410437736.9.

[0003] In addition, deep learning frameworks usually involve complex operations on multi-dimensional tensors. The existing technology is not perfect in supporting multi-dimensional shapes, resulting in frequent shape mismatch problems, which require manual adjustment and are difficult to meet the dynamically changing computing needs, limiting the flexible support for multi-input and multi-shape data. The existing technology usually introduces a tensor broadcasting mechanism to adapt to the computing requirements of multi-shape input tensors. However, during the tensor broadcasting process, the tensor is extended in multiple dimensions, and the system needs to allocate temporary memory for each extended dimension, which not only occupies system memory but also affects computing efficiency. Summary of the Invention

[0004] The technical problem to be solved by the present invention: Aiming at the above problems of the existing technology, a method, system and medium for implementing ONNX deep learning operators on OpenCL devices are provided. The present invention can achieve efficient and general computing support for ONNX operators on OpenCL devices, break through the dependence of deep learning frameworks on a single hardware, reduce the computing overhead during multi-dimensional tensor calculations, and improve the applicability to tensors of different shapes.

[0005] To solve the above technical problems, the technical solution adopted by the present invention is:

[0006] A method for implementing ONNX deep learning operators on OpenCL devices, comprising:

[0007] Step S01, configure the OpenCL ONNX operator. The execution process of the OpenCL ONNX operator includes:

[0008] Step S101, obtain the shape of the input tensor, and determine the shape of the output tensor according to the shape of the input tensor and the type of the target operator to be used;

[0009] Step S102, allocate memory on the OpenCL device according to the shapes of the input tensor and the output tensor;

[0010] Step S103, configure the parameters required by the target operator and pass the parameters to the OpenCL kernel;

[0011] Step S104, convert the index of the one-dimensional linear storage data of the output tensor into a multi-dimensional index according to the shape and stride of the output tensor, establish the mapping relationship between the multi-dimensional index of the output tensor and the multi-dimensional index of the input tensor according to the multi-dimensional index of the output tensor, the type of the target operator, the shape and stride of the input tensor, and then convert the multi-dimensional index of the input tensor into a one-dimensional index to obtain the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor;

[0012] Step S105, calculate the value at each position of the output tensor by using the target operator for the value at each position of each input tensor according to the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor;

[0013] Step S02, execute the OpenCL ONNX operator on the OpenCL device according to the parameters required by the received target operator. During the execution process, perform operator operations on the GPU. After the operator operations are completed, judge whether to transfer to the CPU side or keep in the GPU for the next operator operation according to the execution result.

[0014] Optionally, step S101 includes: starting from the last axis of the input tensor shape for alignment. If the dimensions of the two input tensors on the target axis are the same or the dimension of one of the two input tensors on the target axis is 1, then the dimension of the output tensor on the target axis takes the maximum value of the target axes of the two input tensors.

[0015] Optionally, when allocating memory on the OpenCL device according to the shapes of the input tensor and the output tensor in step S102, the allocated memory size is determined according to the number of data corresponding to the shapes of the input tensor and the output tensor and the size of a single data type.

[0016] Optionally, in step S103, the parameters required for configuring the target operator include: pointers to the input tensor and the output tensor, shape information of the input tensor and the output tensor, strides for index conversion in different dimensions, and any one or more of the auxiliary information, where the auxiliary information includes an operation mode and / or a specified configuration method. In this embodiment, a parameter configuration unit is provided specifically for configuring the parameters required by the operator.

[0017] In step S104, the n-dimensional indices i0, i1, i2, i3,..., i n-1 The calculation formula for converting to a one-dimensional index:

[0018] Index = i0 × s0 + i1 × s1 +... + i n-1 × s n-1 ,

[0019] where, i n-1 is the index of the nth dimension of the input tensor, s0 is the stride of the first dimension of the input tensor, s n-1 is the stride of the nth dimension of the input tensor, and the stride output array (s0, s1,..., s n-1 ) is calculated as:

[0020] s n = s n+1 + d n+1 ,

[0021] where, d n is the number of elements in the nth dimension of the input tensor, s n-1 = 1.

[0022] Optionally, before step S104, it further includes mapping the thread index of the OpenCL device to the element index of the output tensor, so that each thread of the OpenCL device is responsible for calculating one or more output elements.

[0023] Optionally, step S02 includes the following steps:

[0024] Step S201, initialize the OpenCL environment, where the initialization includes selecting a computing device, creating a context, creating a command queue, compiling the OpenCL ONNX operator, and loading the configured OpenCL ONNX operator;

[0025] Step S202, allocate memory on the OpenCL device side for storing input data, output data, and intermediate calculation results;

[0026] Step S203, input the data or model to be processed into the CPU in the form of a tensor;

[0027] Step S204, transfer the tensor data on the CPU side to the OpenCL device through a Memcpy operation;

[0028] Step S205, bind the data transferred to the OpenCL device to the OpenCL kernel;

[0029] Step S206, perform an operator operation on the GPU according to the OpenCL ONNX operator;

[0030] Step S207, after the operator operation is completed, determine whether the current result needs to be transferred to the CPU side. If the current result needs to be transferred to the CPU side, then transfer the current result data back to the CPU from the GPU through a MemcpyToHost operation; if the current result is still needed for subsequent GPU operations, then use the current result as the input for the next operator operation, bind it to the OpenCL Kernel, and then jump to Step S206;

[0031] Optionally, Step S208: When all computing tasks are completed, release the memory allocated on the OpenCL device side.

[0032] In addition, the present invention also provides a system for implementing an ONNX deep learning operator on an OpenCL device, including a microprocessor and a memory connected to each other, and the microprocessor is programmed or configured to execute the method for implementing the ONNX deep learning operator on the OpenCL device.

[0033] In addition, the present invention also provides a computer-readable storage medium, in which a computer program or instruction is stored, and the computer program or instruction is programmed or configured to execute the method for implementing the ONNX deep learning operator on the OpenCL device through a processor.

[0034] In addition, the present invention also provides a computer program product, including a computer program or instruction, and the computer program or instruction is programmed or configured to execute the method for implementing the ONNX deep learning operator on the OpenCL device through a processor.

[0035] Compared with the prior art, the present invention mainly has the following advantages:

[0036] By configuring OpenCL ONNX operators, general computing support on OpenCL devices (such as GPUs, FPGAs, NPUs, etc.) can be achieved, with good scalability. In the execution process of OpenCL ONNX operators, by converting the index of one-dimensional linear storage data of the output tensor into a multi-dimensional index and establishing the mapping relationship between the multi-dimensional index of the output tensor and the multi-dimensional index of the input tensor, the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor can be obtained. When performing calculations, according to the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor, the values at each position of each input tensor are calculated using the target operator to obtain the values at each position of the output tensor, which can simplify the multi-dimensional data access in the broadcast and calculation processes, make the multi-dimensional tensor operations more intuitive and easy to implement, avoid expanding the dimensions of the input tensor during the calculation process, reduce the memory occupation during the calculation process, enable efficient calculation, and can accurately locate the data of the input tensor, ensure that all output elements are correctly calculated and stored in the output tensor, and then can correctly handle the element relationship under the broadcast rule, and have good applicability to the calculation of tensors with different shapes. Brief Description of the Drawings

[0037] Figure 1 It is a flowchart of the execution of the OpenCL ONNX operator of the method of the embodiment of the present invention.

[0038] Figure 2 It is a flowchart of the OpenCL device of the method of the embodiment of the present invention executing the OpenCL ONNX operator. Detailed Embodiments

[0039] Next, the technical solutions of the present invention will be further described in detail in conjunction with the drawings.

[0040] As Figure 1 shown, the implementation method of the ONNX deep learning operator of this embodiment on the OpenCL device includes:

[0041] Step S01, configure the OpenCL ONNX operator, and the execution process of the OpenCL ONNX operator includes:

[0042] Step S101, obtain the shape of the input tensor, and determine the shape of the output tensor according to the shape of the input tensor and the type of the target operator to be used;

[0043] Step S102, allocate memory on the OpenCL device according to the shapes of the input tensor and the output tensor;

[0044] Step S103, configure the parameters required by the target operator and pass the parameters to the OpenCL kernel;

[0045] Step S104: Convert the index of the one-dimensional linear storage data of the output tensor into a multi-dimensional index according to the shape of the output tensor and its stride. Establish the mapping relationship between the multi-dimensional index of the output tensor and the multi-dimensional index of the input tensor according to the multi-dimensional index of the output tensor, the target operator type, the shape of the input tensor and its stride, and then convert the multi-dimensional index of the input tensor into a one-dimensional index to obtain the relationship between the input tensor index and the output tensor index;

[0046] Step S105: Calculate the value at each position of each input tensor using the target operator according to the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor to obtain the value at each position of the output tensor;

[0047] Step S02: Execute the OpenCL ONNX operator on the OpenCL device according to the parameters required by the received target operator. During the execution process, perform the operator operation on the GPU. After the operator operation is completed, judge whether to transfer to the CPU side or remain in the GPU for the next operator operation according to the execution result.

[0048] In this implementation, by configuring the OpenCL ONNX operator, general computing support on OpenCL devices (such as GPUs, FPGAs, NPUs, etc.) can be achieved, and the scalability is good; in the execution process of the OpenCL ONNX operator, by converting the index of the one-dimensional linear storage data of the output tensor into a multi-dimensional index and establishing the mapping relationship between the multi-dimensional index of the output tensor and the multi-dimensional index of the input tensor, the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor can be obtained. When calculating, calculate the value at each position of each input tensor using the target operator according to the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor to obtain the value at each position of the output tensor, which can simplify the multi-dimensional data access in the broadcast and calculation processes, make the multi-dimensional tensor operation more intuitive and easy to implement, avoid expanding the dimensions of the input tensor during the calculation process, reduce the memory occupation during the calculation process, be able to achieve efficient calculation, and be able to accurately locate the data of the input tensor, ensure that all output elements are correctly calculated and stored in the output tensor, and then can correctly handle the element relationship under the broadcast rule, and have good applicability to the calculation of tensors with different shapes.

[0049] Corresponding to Step S105, the calculation formula for the value at the corresponding position of each output tensor is:

[0050] Result data result_index = Calculate operand 1 index1 OP1 Calculate operand 2 index2 OP2 Calculate operand 3(1)

[0051] As an alternative embodiment, step S101 includes: starting from the last axis of the input tensor shape, if the dimensions of the two input tensors on the target axis are the same or the dimension of one of the two input tensors on the target axis is 1, the dimension of the output tensor on the target axis takes the maximum value of the target axis of the two input tensors. For example, some operators (such as addition, multiplication) need to broadcast inputs of different shapes, so the output shape needs to be determined according to the broadcast rules. For example, when using the Add operator to calculate the sum of operand 1 and operand 2, the operand shapes 1(4, 1, 5) and operand 2(1, 3, 5) satisfy shape broadcasting, and the output result shape is (4, 3, 5). Deriving the correct output shape can help with subsequent memory allocation and calculation arrangements.

[0052] As an alternative embodiment, when allocating memory on the OpenCL device according to the shapes of the input tensor and the output tensor in step S102, the allocated memory size is determined according to the number of data corresponding to the shapes of the input tensor and the output tensor and the size of a single data type. For some large tensors or operators that require broadcast operations, allocating sufficient memory space can avoid memory shortage problems during the calculation process and optimize data access efficiency. Specifically, the specific size of the allocated memory is: the number of data represented by the shapes of the input tensor and the output tensor multiplied by the size of a single data type.

[0053] As an alternative embodiment, in step S103, configuring the parameters required for the target operator includes: any one or more of the pointers of the input tensor and the output tensor, the shape information of the input tensor and the output tensor, the stride for index conversion in different dimensions, and auxiliary information. The auxiliary information includes the operation mode and / or the specified configuration method. The functions of the relevant parameters are as follows:

[0054] Pointer of the input / output tensor: used to point to the memory address on the device;

[0055] Shape information: the dimension information of the input tensor and the output tensor;

[0056] Stride: used for index conversion between different dimensions. Assuming the shape of the tensor is (d0, d1, d2, d3, d4), the maximum size n of the data shape array is 5, and in the stride output array (s0, s1, s2, s3, s4),

[0057] The stride corresponding to the last dimension of the tensor: s4 = 1, (2)

[0058] The strides corresponding to other dimensions: s n = s i+1 + d i+1 (3)

[0059] As an alternative embodiment, in step S104, the calculation formula for converting the n-dimensional indices i0, i1, i2, i3, ..., i n-1 into a one-dimensional index Index is:

[0060] Index = i0 × s0 + i1 × s1 +... + i n-1 × s n-1 (4)

[0061] where, i n-1 is the index of the n-th dimension of the input tensor, s0 is the stride of the first dimension of the input tensor, and s n-1 is the stride of the n-th dimension of the input tensor. The calculation formula for the stride output array (s0, s1, ..., s n-1 ) is:

[0062] s n = s n+1 + d n+1 (5)

[0063] where, d n is the number of elements in the n-th dimension of the input tensor. According to the above formula, the multi-dimensional indices of the input tensor can be mapped to a one-dimensional index.

[0064] In step S104, the calculation formula for converting the one-dimensional index of the output tensor into an n-dimensional index is:

[0065]

[0066] As an alternative embodiment, before step S104, it further includes mapping the thread index of the OpenCL device to the element index of the output tensor, so that each thread of the OpenCL device is responsible for calculating one or more output elements. Specifically, it includes converting the thread index obtained by using the built-in function get_global_id() of OpenCL into a multi-dimensional index through dimension expansion calculation, and inversely mapping it to the data storage location of the input tensor, so as to support the index mapping of input tensors with different shapes and different dimensions (N dimensions), enabling the OpenCL computing core to complete data calculation without depending on fixed dimensions, and supporting the parallel execution of operator operations to accelerate the calculation speed.

[0067] The following is the code of a template macro function for an element - level binary calculation operation in OpenCL, which supports 5 - dimensional dynamic input and element - level broadcast features. By defining the binary operation, defining the function body to pass input and output data and necessary shape and stride parameters, configuring the one - dimensional global_id to represent the one - dimensional result data index of the output tensor, using the stride information to calculate the multi - dimensional index indices corresponding to global_id (corresponding to converting the one - dimensional linear storage data index of the output tensor into a multi - dimensional index in this embodiment), determining the corresponding indices of input tensors A and B according to the multi - dimensional index indices (corresponding to establishing the mapping relationship between the multi - dimensional indices of the output tensor and the multi - dimensional indices of the input tensors in this embodiment, and then converting the multi - dimensional indices of the input tensors into one - dimensional indices to obtain the relationship between the one - dimensional indices and the one - dimensional indices of the output tensor), and calculating the output by taking values of data according to the corresponding indices, etc., the operator is implemented in a pipeline - like manner (corresponding to calculating the values of each position of the output tensor using the target operator for the values of each position of each input tensor according to the relationship between the one - dimensional index of the input tensor and the one - dimensional index of the output tensor in this embodiment):

[0068]

[0069]

[0070] It should be noted that in the above code, A and B correspond to the input tensors in this embodiment, C corresponds to the output tensor in this embodiment, Ashapes and Bshapes are the tensor shape arrays of the two input tensors respectively, Cshapes is the tensor shape array of the output tensor, Adim and Bdim are the numbers of dimensions of the two input tensors respectively, and Cdim is the number of dimensions of the output tensor; for different types of operators, the specific steps of determining the source input tensor indices according to the multi - dimensional index indices will also be different.

[0071] As Figure 2 shown, as an optional embodiment, step S02 includes the following steps:

[0072] Step S201, initialize the OpenCL environment, and the initialization includes selecting a computing device, creating a context, creating a command queue, compiling the OpenCL ONNX operator, and loading the configured OpenCL ONNX operator;

[0073] Step S202, allocate memory on the OpenCL device side for storing input data, output data, and intermediate calculation results;

[0074] Step S203, input the data or model to be processed into the CPU in the form of a tensor;

[0075] Step S204, transfer the tensor data on the CPU side to the OpenCL device through the Memcpy operation;

[0076] Step S205, bind the data transferred to the OpenCL device to the OpenCL kernel;

[0077] Step S206, execute the operator operation on the GPU according to the OpenCL ONNX operator;

[0078] Step S207, after the operator operation is completed, determine whether the current result needs to be transferred to the CPU side. If the current result needs to be transferred to the CPU side, then transfer the current result data back to the CPU from the GPU through the MemcpyToHost operation; if the current result still needs to be used for subsequent GPU operations, then use the current result as the input for the next operator operation and bind it to the OpenCL Kernel, and then jump to Step S206;

[0079] Step S208: When all computing tasks are completed, release the memory allocated on the OpenCL device side.

[0080] This step is applicable to scenarios that require frequent CPU-OpenCL device data transfer and multi-operator calculations on the device side to solve complex application scenarios, such as deep learning inference acceleration or large data processing pipelines. Through a flexible data transfer and branch judgment mechanism, the efficiency of CPU-OpenCL device collaborative work can be optimized.

[0081] In addition, this embodiment also provides a system for implementing ONNX deep learning operators on an OpenCL device, including a microprocessor and a memory connected to each other. The microprocessor is programmed or configured to execute the method for implementing ONNX deep learning operators on an OpenCL device.

[0082] In addition, this embodiment also provides a computer-readable storage medium, in which a computer program or instruction is stored. The computer program or instruction is programmed or configured to execute the method for implementing ONNX deep learning operators on an OpenCL device through a processor.

[0083] In addition, this embodiment also provides a computer program product, including a computer program or instruction. The computer program or instruction is programmed or configured to execute the method for implementing ONNX deep learning operators on an OpenCL device through a processor.

[0084] Those skilled in the art should understand that the technical solutions provided by the embodiments of the present application can be in the form of a method, a system, or a computer program product. Therefore, the present application can adopt the form of a complete hardware embodiment, a complete software embodiment, or an embodiment combining software and hardware aspects. Moreover, the present application can adopt the form of a computer program product implemented on one or more computer-readable storage media (including but not limited to disk storage, CD-ROM, optical storage, etc.) that contain computer-usable program code. The present application is described with reference to the flowcharts and / or block diagrams of methods, devices (systems), and computer program products according to the embodiments of the present application. It should be understood that each process and / or block in the flowchart and / or block diagram, as well as the combination of processes and / or blocks in the flowchart and / or block diagram, can be realized by computer program instructions. These computer program instructions can be provided to the processor of a general-purpose computer, a special-purpose computer, an embedded processor, or other programmable data processing devices to generate a machine, so that the instructions executed by the processor of the computer or other programmable data processing devices generate means for realizing the functions specified in the process Figure 1 one process or multiple processes and / or blocks Figure 1 These computer program instructions can also be stored in a computer-readable memory that can direct a computer or other programmable data processing device to work in a specific manner, so that the instructions stored in the computer-readable memory generate a manufactured article including instruction means, and the instruction means realizes the functions specified in the process Figure 1 one process or multiple processes and / or blocks Figure 1 These computer program instructions can also be loaded onto a computer or other programmable data processing device, so that a series of operation steps are executed on the computer or other programmable device to generate a computer-implemented process. Therefore, the instructions executed on the computer or other programmable device provide steps for realizing the functions specified in the process Figure 1 one process or multiple processes and / or blocks Figure 1 one block or multiple blocks.

[0085] The above are only the preferred embodiments of the present invention, and the protection scope of the present invention is not limited to the above embodiments. All technical solutions falling within the concept of the present invention belong to the protection scope of the present invention. It should be noted that for those of ordinary skill in the art, without departing from the principle of the present invention, several improvements and refinements should also be regarded as within the protection scope of the present invention.

Claims

1. A method for implementing an ONNX deep learning operator on an OpenCL device, characterized in that Including: Step S01, configure the OpenCL ONNX operator. The execution process of the OpenCL ONNX operator includes: Step S101, obtain the shape of the input tensor, and determine the shape of the output tensor according to the shape of the input tensor and the type of the target operator to be used; Step S102, allocate memory on the OpenCL device according to the shapes of the input tensor and the output tensor; Step S103, configure the parameters required by the target operator and pass the parameters to the OpenCL kernel; Step S104, convert the one-dimensional linear storage data index of the output tensor into a multi-dimensional index according to the shape and its stride of the output tensor, establish the mapping relationship between the multi-dimensional index of the output tensor and the multi-dimensional index of the input tensor according to the multi-dimensional index of the output tensor, the type of the target operator, the shape and its stride of the input tensor, and then convert the multi-dimensional index of the input tensor into a one-dimensional index to obtain the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor; Step S105, calculate the values at each position of the output tensor using the target operator for the values at each position of each input tensor according to the relationship between the one-dimensional index of the input tensor and the one-dimensional index of the output tensor; Step S02, execute the OpenCL ONNX operator on the OpenCL device according to the parameters required by the received target operator. During the execution process, perform operator operations on the GPU, and after the operator operations are completed, judge whether to transfer to the CPU side or remain in the GPU for the next operator operation according to the execution result.

2. The implementation method of the ONNX deep learning operator according to claim 1 on an OpenCL device, characterized in that, Step S101 includes: starting from the last axis of the input tensor shape for alignment. If the dimensions of the two input tensors on the target axis are the same or the dimension of one of the two input tensors on the target axis is 1, then the dimension of the output tensor on the target axis takes the maximum value of the target axes of the two input tensors.

3. The implementation method of the ONNX deep learning operator according to claim 1 on an OpenCL device, characterized in that, When allocating memory on the OpenCL device according to the shapes of the input tensor and the output tensor in Step S102, the allocated memory size is determined according to the number of data corresponding to the shapes of the input tensor and the output tensor and the size of a single data type.

4. The implementation method of the ONNX deep learning operator according to claim 1 on an OpenCL device, characterized in that, In Step S103, the parameters required to configure the target operator include: any one or more of the pointers of the input tensor and the output tensor, the shape information of the input tensor and the output tensor, the stride for index conversion in different dimensions, and the auxiliary information. The auxiliary information includes the operation mode and / or the specified configuration method.

5. The implementation method of the ONNX deep learning operator according to claim 1 on an OpenCL device, characterized in that, In step S104, the calculation formula for converting the n-dimensional indexes i0, i1, i2, i3,..., i n-1 into a one-dimensional index Index is: Index=i0×s0+i1×s1+...+i n-1 ×s n-1 , where i n-1 is the index of the n-th dimension of the input tensor, s0 is the stride of the first dimension of the input tensor, and s n-1 is the stride of the n-th dimension of the input tensor. The calculation formula for the stride output array (s0, s1,..., s n-1 ) is as follows: s n = s n+1 + d n+1 , where d n is the number of elements in the n-th dimension of the input tensor, and s n-1 = 1.

6. The implementation method of the ONNX deep learning operator according to claim 1 on an OpenCL device, characterized in that, Before Step S104, it also includes mapping the thread index of the OpenCL device to the element index of the output tensor, so that each thread of the OpenCL device is responsible for calculating one or more output elements.

7. The implementation method of the ONNX deep learning operator according to any one of claims 1 to 6 on an OpenCL device, characterized in that, Step S02 includes the following steps: Step S201, initialize the OpenCL environment. The initialization includes selecting a computing device, creating a context, creating a command queue, compiling the OpenCL ONNX operator, and compiling and loading the configured OpenCL ONNX operator; Step S202, allocate memory on the OpenCL device side for storing input data, output data, and intermediate calculation results; Step S203: Input the data or model to be processed in the form of tensors into the CPU; Step S204: Transfer the tensor data on the CPU side to the OpenCL device through the Memcpy operation; Step S205: Bind the data transferred to the OpenCL device to the OpenCL kernel; Step S206: Execute the operator operation on the GPU according to the OpenCL ONNX operator; Step S207: After the operator operation is completed, determine whether the current result needs to be transferred to the CPU side. If the current result needs to be transferred to the CPU side, transfer the current result data from the GPU back to the CPU through the MemcpyToHost operation; if the current result still needs to be used for subsequent GPU operations, use the current result as the input for the next operator operation, bind it to the OpenCL Kernel, and then jump to Step S206; Step S208: When all calculation tasks are completed, release the memory allocated on the OpenCL device side.

8. An implementation system of an ONNX deep learning operator on an OpenCL device, including a microprocessor and a memory connected to each other, characterized in that, The microprocessor is programmed or configured to execute the implementation method of the ONNX deep learning operator according to any one of claims 1 to 7 on the OpenCL device.

9. A computer-readable storage medium storing a computer program or instructions, characterized in that, The computer program or instruction is programmed or configured to execute the implementation method of the ONNX deep learning operator according to any one of claims 1 to 7 on the OpenCL device through the processor.

10. A computer program product, comprising a computer program or instructions, characterized in that, The computer program or instruction is programmed or configured to execute the implementation method of the ONNX deep learning operator according to any one of claims 1 to 7 on the OpenCL device through the processor.

Images