OpenCL TensorFlow implementation method and device, electronic device and storage medium

Abstract

The present disclosure provides an implementation method and device, electronic device and storage medium of OpenCL TensorFlow. Based on the OpenCL programming standard, the OpenCL Device support of the TensorFlow device layer is implemented; with reference to the computing function of the preset kernel in TensorFlow, the OpenCL kernel and OpenCL operator of the corresponding computing function are implemented based on the OpenCL programming standard; based on the OpenCL programming standard, the OpenCLElement‑wise operator library is implemented; the computing acceleration library required for the task is adapted to the OpenCL standard; the API interface of the first OpenCL computing acceleration library is used to implement integrated calls; StreamExecutor's support for OpenCL is implemented, and the OpenCL backend of StreamExecutor is used to implement integrated calls to the second OpenCL computing acceleration library. The present invention implements the OpenCL version of the TensorFlow framework, thereby enabling the TensorFlow framework to support multiple types of heterogeneous devices from different manufacturers.

Description

Technical Field

[0001] The present disclosure relates to the field of artificial intelligence technology, and in particular to an OpenCL TensorFlow implementation method and device, an electronic device, and a storage medium. Background Art

[0002] In recent years, artificial intelligence algorithms and models, exemplified by deep learning, have garnered widespread attention and research. Deep learning has achieved breakthroughs in fields such as image recognition and text processing. At the same time, as the number of model parameters and the amount of data required for training have grown, model training and inference have become increasingly dependent on powerful computing capabilities. To improve the efficiency of deep learning models' use of computing devices and streamline model building and computation, renowned companies and research institutions both domestically and internationally have designed and implemented a variety of machine learning frameworks.

[0003] Currently, machine learning frameworks, such as TensorFlow, generally use a heterogeneous computing architecture combining CPUs and accelerators to improve model computational speed. This architecture leverages the CPU's scheduling and management expertise and the accelerator's advantages in parallelism and single-machine computing peak performance. Compared to traditional homogeneous computing, heterogeneous computing architectures offer higher computing efficiency and lower latency. The high performance and well-developed development ecosystem of GPUs from international manufacturers have led to the official versions of mainstream machine learning frameworks, such as TensorFlow and PyTorch, being primarily developed using closed-source programming models.

[0004] Mainstream machine learning frameworks generally only support computing devices primarily based on GPU chips from foreign manufacturers. The enormous potential of general-purpose accelerators, especially domestic accelerators, in next-generation AI applications is limited. The latest deep learning research results, both domestically and internationally, are largely computed using products from foreign chip companies. The innovation and development of AI technology has become firmly tied to these chip companies. Therefore, ensuring that mainstream deep learning frameworks support multiple types of heterogeneous devices from different manufacturers has become a pressing issue. Summary of the Invention

[0005] This disclosure provides an OpenCL TensorFlow implementation method and device, electronic device, and storage medium. Its primary purpose is to address the issue of TensorFlow supporting multiple types of heterogeneous devices from different manufacturers.

[0006] According to a first aspect of the present disclosure, a method for implementing OpenCL TensorFlow is provided, comprising:

[0007] Based on the OpenCL programming standard, OpenCL Device support is implemented for the TensorFlow device layer;

[0008] Refer to the calculation function of the preset kernel in TensorFlow and implement the OpenCL kernel and OpenCL operator with the corresponding calculation function based on the OpenCL programming standard;

[0009] Implement the OpenCL Element-wise operator library based on the OpenCL programming standard;

[0010] Adapt the computing acceleration library required for the task to the OpenCL standard;

[0011] Using the API interface of the first OpenCL computing acceleration library to implement integrated calls;

[0012] Implement StreamExecutor's support for OpenCL and use StreamExecutor's OpenCL backend to implement integrated calls to the second OpenCL computing acceleration library.

[0013] Optionally, the OpenCL Device support for implementing the TensorFlow device layer based on the OpenCL programming standard includes:

[0014] The classes with different functions of OpenCL Device are implemented by inheritance, and the functions of obtaining devices and related contexts during the calculation process of various operators are implemented by using the classes with different functions.

[0015] Optionally, referring to the computing function of the preset kernel in TensorFlow, the OpenCL kernel and OpenCL operator implementing the corresponding computing function based on the OpenCL programming standard include:

[0016] Encapsulate the OpenCL kernel into an OpenCL operator and register the OpenCL operator in TensorFlow.

[0017] Optionally, the implementation of the OpenCL Element-wise operator library based on the OpenCL programming standard includes:

[0018] Implement the underlying OpenCL kernels of the Element-Wise operators one by one;

[0019] The underlying OpenCL kernel is encapsulated to generate an OpenCL operator.

[0020] Optionally, adapting the computing acceleration library required for the task to the OpenCL standard includes:

[0021] Implement the corresponding underlying OpenCL kernel of the computing acceleration library's operators, encapsulate and generate the corresponding OpenCL operators, and register them in TensorFlow.

[0022] Optionally, implementing StreamExecutor's support for OpenCL and utilizing the OpenCL backend of StreamExecutor to implement integrated calls to a second OpenCL computing acceleration library includes:

[0023] Implement the OpenCL Platform in StreamExecutor and register it with the TensorFlow framework;

[0024] The OpenCL Platform is created according to the instructions and specifies the corresponding device Platform;

[0025] Implement OpenCL Executor in StreamExecutor and register it with the TensorFlow framework;

[0026] Complete the registration and access of the computing acceleration library in the LibrarySupport layer of StreamExecutor;

[0027] In the TensorFlow framework, bind the corresponding operators in the computing acceleration library to the OpenCL device and register them.

[0028] Optionally, the method further includes:

[0029] Optimize the implementation process of OpenCL Kernel and OpenCL operators.

[0030] Optionally, the optimizing the implementation process of the OpenCL Kernel and the OpenCL operator includes:

[0031] Use the singleton pattern to initialize the OpenCL platform, device, context, and command queue;

[0032] Cache the compilation results of OpenCL Kernel programs;

[0033] Use reduction algorithms to optimize atomic operations.

[0034] According to a second aspect of the present disclosure, there is provided an OpenCL TensorFlow implementation apparatus, comprising:

[0035] The first generation unit is used to implement OpenCL Device support for the TensorFlow device layer;

[0036] The second generation unit is used to refer to the computing function of the preset kernel in TensorFlow and implement the OpenCL kernel and OpenCL operator of the corresponding computing function based on the OpenCL programming standard;

[0037] The third generation unit is used to implement the OpenCL Element-wi se operator library based on the OpenCL standard;

[0038] The adaptation unit is used to adapt the computing acceleration library required by the task to the OpenCL standard;

[0039] A first calling unit is configured to implement an integrated call using an API interface of a first OpenCL computing acceleration library;

[0040] The second calling unit is used to implement StreamExecutor's support for OpenCL and use StreamExecutor's OpenCL backend to implement integrated calls to the OpenCL computing acceleration library.

[0041] Optionally, the first generating unit is further configured to:

[0042] The classes with different functions of OpenCL Device are implemented by inheritance, and the functions of obtaining devices and related contexts during the calculation process of various operators are implemented by using the classes with different functions.

[0043] Optionally, the second generating unit is further configured to:

[0044] Encapsulate the OpenCL kernel into an OpenCL operator and register the OpenCL operator in TensorFlow.

[0045] Optionally, the third generating unit includes:

[0046] A first generation module is used to implement the underlying OpenCL kernel of the Element-Wise operator one by one;

[0047] The encapsulation module is used to encapsulate the device-side code and functions of the underlying OpenCL kernel to generate an OpenCL operator.

[0048] Optionally, the adaptation unit is further configured to:

[0049] Implement the corresponding underlying OpenCL kernel of the computing acceleration library's operators, encapsulate and generate the corresponding OpenCL operators, and register them in TensorFlow.

[0050] Optionally, the second calling unit includes:

[0051] The second generation module is used to implement the OpenCL Platform in StreamExecutor and register it with the TensorFlow framework;

[0052] A creation module is used to create the OpenCL Platform according to the instruction and specify the corresponding device Platform;

[0053] The third generation module is used to implement OpenCL Executor in StreamExecutor and register it with the TensorFlow framework;

[0054] The registration module is used to complete the registration and access of the computing acceleration library in the LibrarySupport layer of StreamExecutor;

[0055] The binding module is used to bind and register the corresponding operators in the computing acceleration library with OpenCL devices in the TensorFlow framework.

[0056] Optionally, the device further includes:

[0057] The optimization unit is used to optimize the implementation process of OpenCL kernels and OpenCL operators.

[0058] Optionally, the optimization unit includes:

[0059] Initialization module, used to implement the initialization of OpenCL platform, device, context and command queue OpenCL using singleton mode;

[0060] Cache module, used to cache the compilation results of OpenCL kernel programs;

[0061] The optimization module is used to optimize atomic operations using reduction algorithms.

[0062] According to a third aspect of the present disclosure, there is provided an electronic device, including:

[0063] at least one processor; and

[0064] a memory communicatively connected to the at least one processor; wherein,

[0065] The memory stores instructions that can be executed by the at least one processor. The instructions are executed by the at least one processor to enable the at least one processor to perform the method described in the first aspect.

[0066] According to a fourth aspect of the present disclosure, a non-transitory computer-readable storage medium storing computer instructions is provided, wherein the computer instructions are used to enable the computer to execute the method described in the first aspect.

[0067] According to a fifth aspect of the present disclosure, a computer program product is provided, comprising a computer program, wherein when the computer program is executed by a processor, the computer program implements the method as described in the first aspect above.

[0068] The present disclosure provides an implementation method and device, electronic device and storage medium of OpenCL TensorFlow. Based on the OpenCL programming standard, the OpenCL Device support of the TensorFlow device layer is implemented; with reference to the computing function of the preset kernel in TensorFlow, the OpenCL kernel and OpenCL operator of the corresponding computing function are implemented based on the OpenCL programming standard; based on the OpenCL programming standard, the OpenCL Element-wise operator library is implemented; the computing acceleration library required for the task is adapted to the OpenCL standard; the API interface of the first OpenCL computing acceleration library is used to implement integrated calls; StreamExecutor's support for OpenCL is implemented, and the OpenCL backend of StreamExecutor is used to implement integrated calls to the second OpenCL computing acceleration library. The present invention implements the OpenCL version of the TensorFlow framework, thereby enabling the TensorFlow framework to support multiple types of heterogeneous devices from different manufacturers.

[0069] It should be understood that the content described in this section is not intended to identify the key or important features of the embodiments of the present application, nor is it intended to limit the scope of the present application. Other features of the present application will become easily understood through the following description. BRIEF DESCRIPTION OF THE DRAWINGS

[0070] The accompanying drawings are provided to facilitate a better understanding of the present invention and do not constitute a limitation of the present disclosure.

[0071] Figure 1 A flowchart of an implementation method of OpenCL TensorFlow provided in an embodiment of the present disclosure;

[0072] Figure 2 A statistical diagram of the proportion of OpenCL interface execution time in a typical OpenCL kernel code provided in an embodiment of the present disclosure;

[0073] Figure 3 A schematic diagram of a TensorFlow singleton class provided in an embodiment of the present disclosure;

[0074] Figure 4 A schematic diagram of multi-workgroup atomic operation calculation in TensorFlow provided in an embodiment of the present disclosure;

[0075] Figure 5 A two-step reduction calculation diagram provided for an embodiment of the present disclosure

[0076] Figure 6 A schematic diagram of the structure of an OpenCL TensorFlow implementation device provided in an embodiment of the present disclosure;

[0077] Figure 7 A schematic diagram of the structure of another OpenCL TensorFlow implementation device provided in an embodiment of the present disclosure;

[0078] Figure 8 A schematic block diagram of an exemplary electronic device 300 provided in accordance with an embodiment of the present disclosure. DETAILED DESCRIPTION

[0079] The following description of exemplary embodiments of the present disclosure is made in conjunction with the accompanying drawings, including various details of the embodiments of the present disclosure to facilitate understanding. These details should be considered as merely exemplary. Therefore, those skilled in the art will recognize that various changes and modifications may be made to the embodiments described herein without departing from the scope and spirit of the present disclosure. Similarly, for the sake of clarity and conciseness, descriptions of well-known functions and structures are omitted in the following description.

[0080] The following describes the OpenCL TensorFlow implementation method and apparatus, electronic device, and storage medium according to the embodiments of the present disclosure with reference to the accompanying drawings.

[0081] Figure 1 A flowchart of an OpenCL TensorFlow implementation method provided in an embodiment of the present disclosure.

[0082] like Figure 1 As shown, the method comprises the following steps:

[0083] Step 101: Implement OpenCL Device support for the TensorFlow device layer based on the OpenCL programming standard.

[0084] OpenCL Device support mainly requires completing memory allocation and release, Tensor lifecycle management, data interaction support, kernel execution support, and management of objects or attributes related to storage devices.

[0085] Step 102 : referring to the computing function of the preset kernel in TensorFlow and based on the OpenCL programming standard, implement the OpenCL kernel and OpenCL operator of the corresponding computing function.

[0086] Refer to the computing functions of the kernel in the existing programming model in TensorFlow to implement the corresponding OpenCL kernel and complete the parallel implementation of the computing functions on the OpenCL device.

[0087] Step 103: Implement the OpenCL Element-wise operator library based on the OpenCL programming standard.

[0088] Element-Wise operators, a key component of the TensorFlow framework, are widely used and implement a wide range of computing functions. Each Element-Wise operator is implemented as an underlying OpenCL kernel and registered with the deep learning framework.

[0089] Step 104: Adapt the computing acceleration library required for the task to the OpenCL standard.

[0090] According to the requirements of different tasks, different computing acceleration libraries can be integrated into the TensorFlow framework. In the embodiment of the present disclosure, the computing acceleration library also needs to be adapted to the OpenCL standard.

[0091] Step 105: Implement integrated calling using the API interface of the first OpenCL computing acceleration library.

[0092] In the TensorFlow framework, the integrated call of the computing acceleration library can be based on the StreamExecutor call, or it can be independently realized by directly calling the acceleration library API interface. The first OpenCL computing acceleration library is an OpenCL computing acceleration library that is directly called.

[0093] Step 106 : Implement StreamExecutor's support for OpenCL, and use the OpenCL backend of StreamExecutor to implement integrated calls to the second OpenCL computing acceleration library.

[0094] StreamExecutor's OpenCL backend implements integrated calls to OpenCL compute acceleration libraries, including the OpenCL DNN library, the OpenCL BLAS library, and the OpenCL high-performance computing library. This integrated call to different compute acceleration libraries can improve the efficiency of deep learning frameworks in specific operations. The second OpenCL compute acceleration library is the one called by StreamExecutor.

[0095] The present disclosure provides an implementation method of OpenCL TensorFlow. Based on the OpenCL programming standard, the OpenCL Device support of the TensorFlow device layer is implemented; with reference to the computing function of the preset kernel in TensorFlow, the OpenCL kernel and OpenCL operator of the corresponding computing function are implemented based on the OpenCL programming standard; based on the OpenCL programming standard, the OpenCL Element-wise operator library is implemented; the computing acceleration library required for the task is adapted to the OpenCL standard; the API interface of the first OpenCL computing acceleration library is used to implement integrated calls; StreamExecutor's support for OpenCL is implemented, and the OpenCL backend of StreamExecutor is used to implement integrated calls to the second OpenCL computing acceleration library. The present invention implements the OpenCL version of the TensorFlow framework, thereby enabling the TensorFlow framework to support multiple types of heterogeneous devices from different manufacturers.

[0096] In an embodiment of the present application, an OpenCL version of the TensorFlow framework ncl-TensorFlow is constructed according to the aforementioned method.

[0097] As a feasible method of the embodiment of the present application, the following method may be adopted but is not limited to:

[0098] Furthermore, in the embodiment of the present application, the OpenCL Device support for the TensorFlow device layer based on the OpenCL programming standard includes:

[0099] The classes with different functions of OpenCL Device are implemented by inheritance, and the functions of obtaining devices and related contexts during the calculation process of various operators are implemented by using the classes with different functions.

[0100] Based on the type inheritance rules in the TensorFlow framework, the allocator, device_context, device_factory, and device classes of the OpenCL device are inherited and implemented to support the functions of obtaining devices and related contexts in the calculation process of various operators in the TensorFlow framework.

[0101] Implement memory allocation and release support for OpenCL devices: Based on the inheritance of the Allocator class in the TensorFlow framework, the memory allocation and release functions of OpenCL devices are implemented. The Tensor creation and release functions based on OpenCL devices are implemented.

[0102] Data exchange and management between the host and OpenCL devices: This class inherits the DeviceContext class from the TensorFlow framework, implementing Tensor copying and Tensor lifecycle management between the OpenCL device and host. By implementing the DeviceContext class, users no longer need to explicitly copy and manage data storage locations or Tensors; DeviceContext manages the Tensor lifecycle and automatically releases memory.

[0103] Support for OpenCL devices: This class inherits from the LocalDevice class in the TensorFlow framework and implements the OpenCL Device class. It binds Allocator and DeviceContext objects to implement memory allocation, Tensor creation, and memory management. It also inherits from the DeviceFactory class in the TensorFlow framework to create OpenCL devices using the factory pattern. The OpenCL Device class implements memory allocation and management, as well as storage of OpenCL device information and properties, such as memory limits, device name, maximum workgroup size, and shared memory size.

[0104] Furthermore, in the embodiment of the present application, referring to the computing function of the preset kernel in TensorFlow, the OpenCL kernel and OpenCL operator implementing the corresponding computing function based on the OpenCL programming standard include:

[0105] Encapsulate the OpenCL kernel into an OpenCL operator and register the OpenCL operator in TensorFlow.

[0106] The CUDA kernel in the TensorFlow source code contains a large number of operators with diverse calculation types and functions. Summarizing a set of standard conversion processes can greatly improve conversion efficiency and code quality, and reduce coding errors during the conversion process. This application embodiment is based on the research and analysis of the implementation characteristics of the CUDA kernel in the TensorFlow source code, and summarizes the conversion process from the CUDA kernel in TensorFlow to the OpenCL kernel in general into the following three steps:

[0107] The C++ code implementation in the kernel is replaced with the code implementation that complies with the OpenCL C99 standard;

[0108] The CUDA kernel in TensorFlow is developed in C++, while the kernel code in the OpenCL 1.2 standard only supports the C99 standard. Therefore, operators such as reinterpret_cast() in C++ need to be replaced. For some CUDA kernels that include structures or classes as input parameters, the members of the structure or class need to be declared as input parameters of the OpenCL kernel, and the methods in the class need to be rewritten as sub-functions that can be directly called by the OpenCL kernel.

[0109] Kernel modifier and interface function replacement;

[0110] During the kernel conversion process, CUDA kernel modifiers and interface functions need to be replaced based on the correspondence between CUDA and OpenCL basic concepts in Table 1. For example, the __global__ qualifier in the CUDA kernel is modified to the kernel in OpenCL; and the syncthreads() function used for synchronization in the CUDA kernel is replaced with the barrier() function in OpenCL.

[0111] Unlike typical CUDA kernel code, the CUDA kernel in the TensorFlow source code typically contains wrappers for built-in CUDA variables or interfaces. Furthermore, some host-side interface functions in the CUDA code must be modified. For example, the cudaLaunchKernel() function in the CUDA host code, used to send the kernel to the GPU, must be replaced with the OpenCL clEnqueueNDRangeKernel() function.

[0112] Table 1

[0113]

[0114] Rewrite the device function code in Kernel;

[0115] The CUDA kernel often abstracts some computing functionality into functions qualified by the "__device__" modifier and calls them from within the CUDA kernel. Since there is no corresponding modifier in the OpenCL standard, this article reimplements a sub-function with the same functionality during the conversion process and calls it from within the OpenCL kernel.

[0116] Furthermore, in the embodiment of the present application, the implementation of the OpenCL Element-wise operator library based on the OpenCL programming standard includes:

[0117] Implement the underlying OpenCL kernels of the Element-Wise operators one by one;

[0118] The underlying OpenCL kernel is encapsulated to generate an OpenCL operator.

[0119] Based on the commonality of operators of Element-Wise operators, two general operator classes are generated; the computational functions in the two operator classes are encapsulated through inheritance; based on the OpenCL standard, the Element-Wise operators are implemented one by one as underlying OpenCL kernels; and the device-side code and functions of the underlying OpenCL kernels are encapsulated.

[0120] After analyzing the execution characteristics of OpenCL, the base structure was rewritten. The get_kernel_src function was used to implement the corresponding OpenCL kernel string generation function in different subclasses. Based on the input and output data types passed in, it generates OpenCL device-side code strings that complete the corresponding functions and use different data types.

[0121] Furthermore, in the embodiment of the present application, adapting the computing acceleration library required by the task to the OpenCL standard includes:

[0122] Implement the corresponding underlying OpenCL kernel of the computing acceleration library's operators, encapsulate and generate the corresponding OpenCL operators, and register them in TensorFlow.

[0123] The TensorFlow framework can integrate various acceleration libraries to improve computational efficiency in specific applications. However, the ncl-TensorFlow framework requires OpenCL adaptation for each acceleration library.

[0124] Adaptation of the OpenCL DNN library (oclDNN) and the OpenCL high-performance computing library (oclCUB). The oclDNN library implements common operators in deep neural networks, such as convolution and pooling; the oclCUB library implements parallel implementation of common algorithms, as well as sorting, scanning, and reduction operations for OpenCL multithreading.

[0125] Furthermore, in the embodiment of the present application, the implementation of StreamExecutor's support for OpenCL and the use of StreamExecutor's OpenCL backend to implement integrated calls to the second OpenCL computing acceleration library include:

[0126] Implement the OpenCL Platform in StreamExecutor and register it with the TensorFlow framework; the OpenCL Platform is created according to the instructions and specifies the corresponding device Platform;

[0127] Implement OpenCL Executor in StreamExecutor and register it with the TensorFlow framework;

[0128] Complete the registration and access of the computing acceleration library in the LibrarySupport layer of StreamExecutor;

[0129] In the TensorFlow framework, bind the corresponding operators in the computing acceleration library to the OpenCL device and register them.

[0130] The Executor layer hides the complex design of the Platform and provides a unified StreamExecutor class for upper-layer calls. The LibrarySupport layer provides access to the underlying libraries used for machine learning calculations. By registering and connecting the acceleration library to LibrarySupport, StreamExecutor can call the acceleration library for accelerated calculations during calculations.

[0131] Furthermore, in an embodiment of the present application, the method further includes:

[0132] Optimize the implementation process of OpenCL kernel and OpenCL operators.

[0133] Through actual testing, we found that the performance of the converted OpenCL kernel could be further optimized and improved. Therefore, we specifically studied the characteristics of OpenCL code during device initialization, compilation and execution, and designed and implemented three optimization methods.

[0134] Furthermore, in the embodiment of the present application, the optimization of the implementation process of the OpenCL kernel and the OpenCL operator includes:

[0135] Use the singleton pattern to initialize the OpenCL platform, device, context, and command queue;

[0136] Cache the compilation results of OpenCL kernel programs;

[0137] Use reduction algorithms to optimize atomic operations.

[0138] Use the singleton pattern to initialize the OpenCL platform, device, context, and command queue;

[0139] When the host code of an OpenCL program is initialized, it uses API functions such as clCreateContext() and clCreateCommandQueue() to perform operations such as searching for the platform and device, creating a context, and creating a command queue. After the OpenCL kernel operation is completed, the device, context, command queue, and other resources requested during initialization must be released by calling clRelease().

[0140] Figure 2 This chart shows the time spent in OpenCL interface functions during a single execution of six frequently used OpenCL kernels in machine learning models. The chart clearly shows that OpenCL initialization and release functions, such as clCreateContext(), clCreateCommandQueue(), and clRelease(), collectively account for approximately 90% of the code execution time. In machine learning models, a single operation often requires the simultaneous execution of multiple kernels. This means that the relevant OpenCL interface functions are repeatedly called, significantly reducing the execution efficiency of the entire model.

[0141] In fact, according to the OpenCL standard, initialization-related interface functions such as platform search, device application, and context creation only need to be called once when running different kernels. Therefore, this article uses the singleton pattern in C++ to encapsulate these interface functions. Figure 3This article presents the class diagrams designed for some OpenCL interface functions. When executing OpenCL code, if the singleton object hasn't been created, a new one is created and its member variables are initialized. If the singleton object has already been created, the existing object is returned. For operations like initializing a device, creating a context, or creating a command queue, the GetDevice(), GetContext(), and GetQueue() interfaces are used to retrieve initialized variables such as the device, context, and command queue.

[0142] Cache the compiled results of the OpenCL kernel program. The basic process for executing an OpenCL kernel is as follows: First, call the clCreateProgramWithSource() function to compile the OpenCL kernel string into a cl_program object. Second, pass the cl_program object as a parameter to the clCreateKernel() function to generate a cl_kernel object. Finally, send the cl_kernel to the device for execution. As can be seen from the above process, if an OpenCL program calls a kernel multiple times, the program will repeatedly compile the kernel, seriously affecting the execution efficiency of the OpenCL code.

[0143] The cl_program object generated by compiling the OpenCL kernel source code can be reused. A single cl_program object can be used to generate multiple cl_kernel objects. Figure 3 In the Singleton class shown, this article constructs an unordered_map<string,cl_program> The variable program_record of type cl_program is used to cache the cl_program generated by compiling the kernel source code, using the kernel string as an index. The program first checks whether a cl_program corresponding to the kernel string to be called exists in the cache. If so, it is directly retrieved and used. If not, the kernel string is compiled and the generated cl_program is cached.

[0144] Caching OpenCL kernel compilation results can reduce performance loss when compiling the same OpenCL kernel, improving overall program efficiency. Other OpenCL-based machine learning frameworks or math libraries can also be designed or optimized using the same approach.

[0145] Use reduction algorithms to optimize atomic operations;

[0146] When the kernel needs to call atomic operations multiple times to modify the same memory address, a series of atomic operations can only be completed serially due to the competition between processes. Therefore, the computing performance of using atomic operations is sometimes even worse than that of a single-threaded loop. Figure 4 As shown in FIG, although the parallel completion of atomic operations by multiple work groups in OpenCL can improve computing performance to a certain extent, the performance loss caused by process blocking cannot be ignored when the input data volume is large.

[0147] In order to fully utilize the advantages of OpenCL multi-threaded processing of computing tasks and further improve the performance of the converted OpenCL kernel, this paper specifically studies how to use the reduction idea to optimize the OpenCL atomic operation method. Figure 5 As shown in Figure 2, given that OpenCL only supports intra-workgroup thread synchronization, this paper uses a two-stage reduction algorithm to calculate the reduction results: First, the intra-workgroup reduction is performed and the reduction results for each group are saved. Second, a single workgroup is used to perform a further reduction on all the results from the first step to obtain the final result.

[0148] The key to the correct operation of the reduction algorithm lies in the setting of the global worker thread (global_work_size) and local worker thread (local_work_size) sizes in the OpenCL kernel, as well as the calculation of the space required to store intermediate results. In this article, the number of local worker threads is set to a power of 2, and the number of global worker threads and the size of the space required to store the intermediate results of the first step are adjusted based on the input data.

[0149] Corresponding to the aforementioned OpenCL TensorFlow implementation method, the present invention also provides an OpenCL TensorFlow implementation device. Since the device embodiment of the present invention corresponds to the aforementioned method embodiment, any details not disclosed in the device embodiment can be referred to the aforementioned method embodiment and will not be further described in this invention.

[0150] Figure 6 A schematic diagram of the structure of an OpenCL TensorFlow implementation device provided in an embodiment of the present disclosure is shown in FIG. Figure 6 As shown, it includes: a first generating unit 21, a second generating unit 22, a third generating unit 23, an adapting unit 24, a first calling unit 25, and a second calling unit 26.

[0151] A first generating unit 21 is configured to implement OpenCL Device support for the TensorFlow device layer;

[0152] The second generating unit 22 is configured to implement an OpenCL kernel and an OpenCL operator corresponding to the computing function based on the OpenCL programming standard, with reference to the computing function of the preset kernel in TensorFlow;

[0153] The third generating unit 23 is configured to implement an OpenCL Element-wise operator library based on the OpenCL standard;

[0154] The adaptation unit 24 is used to adapt the computing acceleration library required by the task to the OpenCL standard;

[0155] A first calling unit 25 is configured to implement an integrated call using an API interface of a first OpenCL computing acceleration library;

[0156] The second calling unit 26 is used to implement StreamExecutor's support for OpenCL and use the OpenCL backend of StreamExecutor to implement integrated calling of the second OpenCL computing acceleration library.

[0157] Furthermore, in the embodiment of the present disclosure, the first generating unit 21 is further configured to:

[0158] The classes with different functions of OpenCL Device are implemented by inheritance, and the functions of obtaining devices and related contexts during the calculation process of various operators are implemented by using the classes with different functions.

[0159] Furthermore, in the embodiment of the present disclosure, the second generating unit 22 is further configured to:

[0160] Encapsulate the OpenCL kernel into an OpenCL operator and register the OpenCL operator in TensorFlow.

[0161] Furthermore, in the embodiment of the present disclosure, if Figure 7 As shown, the third generating unit 23 includes:

[0162] A first generating module 231 is configured to implement the underlying OpenCL kernels of the Element-Wise operators one by one;

[0163] The encapsulation module 232 is configured to encapsulate the device-side code and functions of the underlying OpenCL kernel to generate an OpenCL operator.

[0164] Furthermore, in the embodiment of the present disclosure, the adaptation unit 24 is further configured to:

[0165] Implement the corresponding underlying OpenCL kernel of the computing acceleration library's operators, encapsulate and generate the corresponding OpenCL operators, and register them in TensorFlow.

[0166] Furthermore, in the embodiment of the present disclosure, if Figure 7 As shown, the second calling unit 26 includes:

[0167] The second generation module 261 is used to implement the OpenCL Platform in the StreamExecutor and register it in the TensorFlow framework;

[0168] A creation module 262 is used to create the OpenCL Platform according to the instruction and specify the corresponding device Platform;

[0169] The third generation module 263 is used to implement the OpenCL Executor in the StreamExecutor and register it in the TensorFlow framework;

[0170] The registration module 264 is used to complete the registration and access of the computing acceleration library in the LibrarySupport layer in the StreamExecutor;

[0171] The binding module 265 is used to bind and register the corresponding operator in the computing acceleration library with the OpenCL device in the TensorFlow framework.

[0172] Furthermore, in the embodiment of the present disclosure, if Figure 7 As shown, the device also includes:

[0173] The optimization unit 27 is used to optimize the implementation process of the OpenCL kernel and the OpenCL operator.

[0174] Furthermore, in the embodiment of the present disclosure, if Figure 7 As shown, the optimization unit 27 includes:

[0175] Initialization module 271, used to implement initialization of OpenCL platform, device, context and command queue OpenCL using singleton mode;

[0176] A cache module 272 is used to cache the compilation results of the OpenCL kernel program;

[0177] The optimization module 273 is used to optimize atomic operations using a reduction algorithm.

[0178] It should be noted that the above explanation of the method embodiment is also applicable to the device of this embodiment, and the principles are the same, which is not limited in this embodiment.

[0179] According to an embodiment of the present disclosure, the present disclosure also provides an electronic device, a readable storage medium, and a computer program product.

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

[0181] like Figure 8 As shown, the device 300 includes a computing unit 301, which can perform various appropriate actions and processes according to a computer program stored in a ROM (Read-Only Memory) 302 or a computer program loaded from a storage unit 308 into a RAM (Random Access Memory) 303. Various programs and data required for the operation of the device 300 can also be stored in the RAM 303. The computing unit 301, ROM 302, and RAM 303 are connected to each other via a bus 304. An I / O (Input / Output) interface 305 is also connected to the bus 304.

[0182] Multiple components in device 300 are connected to I / O interface 305, including: input unit 303, such as a keyboard, mouse, etc.; output unit 307, such as various types of displays, speakers, etc.; storage unit 308, such as a magnetic disk, optical disk, etc.; and communication unit 309, such as a network card, modem, wireless communication transceiver, etc. Communication unit 309 allows device 300 to exchange information / data with other devices via a computer network such as the Internet and / or various telecommunication networks.

[0183] The computing unit 301 can be a variety of general-purpose and / or specialized processing components with processing and computing capabilities. Some examples of the computing unit 301 include, but are not limited to, a CPU (Central Processing Unit), a GPU (Graphic Processing Unit), various specialized AI (Artificial Intelligence) computing chips, various computing units that run machine learning model algorithms, a DSP (Digital Signal Processor), and any suitable processor, controller, microcontroller, etc. The computing unit 301 performs the various methods and processes described above, such as the OpenCL TensorFlow implementation. For example, in some embodiments, the OpenCL TensorFlow implementation can be implemented as a computer software program tangibly embodied in a machine-readable medium, such as the storage unit 308. In some embodiments, part or all of the computer program can be loaded and / or installed on the device 300 via the ROM 302 and / or the communication unit 309. When the computer program is loaded into the RAM 303 and executed by the computing unit 301, one or more steps of the method described above can be performed. Alternatively, in other embodiments, the computing unit 301 may be configured to execute the aforementioned OpenCL TensorFlow implementation method in any other appropriate manner (for example, by means of firmware).

[0184] Various embodiments of the systems and techniques described herein can be implemented in digital electronic circuit systems, integrated circuit systems, FPGAs (Field Programmable Gate Arrays), ASICs (Application-Specific Integrated Circuits), ASSPs (Application-Specific Standard Products), SOCs (System on Chips), CPLDs (Complex Programmable Logic Devices), computer hardware, firmware, software, and / or combinations thereof. These various embodiments can include being implemented in one or more computer programs that are executable and / or interpreted on a programmable system that includes at least one programmable processor, which can be a special-purpose or general-purpose programmable processor that can receive data and instructions from a storage system, at least one input device, and at least one output device, and transmit data and instructions to the storage system, the at least one input device, and the at least one output device.

[0185] The program code for implementing the method of the present disclosure can be written in any combination of one or more programming languages. These program codes can be provided to a processor or controller of a general-purpose computer, a special-purpose computer, or other programmable data processing device so that when the program code is executed by the processor or controller, the functions / operations specified in the flow chart and / or block diagram are implemented. The program code can be executed entirely on the machine, partially on the machine, as a stand-alone software package, partially on the machine and partially on a remote machine, or entirely on a remote machine or server.

[0186] In the context of the present disclosure, a machine-readable medium may be a tangible medium that may contain or store a program for use by or in conjunction with an instruction execution system, device, or apparatus. A machine-readable medium may be a machine-readable signal medium or a machine-readable storage medium. A machine-readable medium may include, but is not limited to, an electronic, magnetic, optical, electromagnetic, infrared, or semiconductor system, device, or apparatus, or any suitable combination of the foregoing. More specific examples of machine-readable storage media may include an electrical connection based on one or more wires, a portable computer disk, a hard disk, RAM, ROM, EPROM (Electrically Programmable Read-Only-Memory) or flash memory, optical fiber, CD-ROM (Compact Disc Read-Only Memory), an optical storage device, a magnetic storage device, or any suitable combination of the foregoing.

[0187] To provide interaction with a user, the systems and techniques described herein can be implemented on a computer having: a display device (e.g., a CRT (Cathode-Ray Tube) or LCD (Liquid Crystal Display) monitor) for displaying information to the user; and a keyboard and pointing device (e.g., a mouse or trackball) through which the user can provide input to the computer. Other types of devices can also be used to provide interaction with the user; for example, the feedback provided to the user can be any form of sensory feedback (e.g., visual feedback, auditory feedback, or tactile feedback); and input from the user can be received in any form (including acoustic input, voice input, or tactile input).

[0188] The systems and techniques described herein can be implemented in a computing system that includes backend components (e.g., as a data server), or a computing system that includes middleware components (e.g., an application server), or a computing system that includes frontend components (e.g., a user computer with a graphical user interface or web browser through which a user can interact with implementations of the systems and techniques described herein), or a computing system that includes any combination of such backend components, middleware components, or frontend components. The components of the system can be interconnected by any form or medium of digital data communication (e.g., a communication network). Examples of communication networks include: LAN (Local Area Network), WAN (Wide Area Network), the Internet, and blockchain networks.

[0189] A computer system may include a client and a server. The client and server are generally remote from each other and typically interact via a communication network. This client-server relationship is established by computer programs running on the respective computers, establishing a client-server relationship. The server may be a cloud server, also known as a cloud computing server or cloud host, a host product within the cloud computing service ecosystem that addresses the management difficulties and limited scalability of traditional physical hosts and VPS services ("Virtual Private Servers" or simply "VPS"). The server may also be a server in a distributed system or a server integrated with blockchain.

[0190] It's important to note that artificial intelligence (AI) is the study of how computers can simulate certain human thought processes and intelligent behaviors (such as learning, reasoning, thinking, and planning). This encompasses both hardware and software technologies. AI hardware technologies generally include sensors, specialized AI chips, cloud computing, distributed storage, and big data processing. AI software technologies primarily encompass computer vision, speech recognition, natural language processing, machine learning / deep learning, big data processing, and knowledge graphs.

[0191] It should be understood that the various forms of the processes shown above can be used to reorder, add, or delete steps. For example, the steps described in this disclosure can be performed in parallel, sequentially, or in a different order, as long as the desired results of the technical solutions disclosed in this disclosure can be achieved. This is not a limitation herein.

[0192] The above specific embodiments do not constitute a limitation on the scope of protection of this disclosure. Those skilled in the art will appreciate that various modifications, combinations, sub-combinations, and substitutions may be made based on design requirements and other factors. Any modifications, equivalent substitutions, and improvements made within the spirit and principles of this disclosure shall be included within the scope of protection of this disclosure.

Claims

1. A method for implementing OpenCL TensorFlow, characterized in that: include: Based on the OpenCL programming standard, OpenCLDevice support for the TensorFlow device layer is implemented; Refer to the calculation function of the preset kernel in TensorFlow, and implement the OpenCL kernel and OpenCL operators with corresponding calculation functions based on the OpenCL programming standard; Implement the OpenCL Element-wise operator library based on the OpenCL programming standard; Adapt the computing acceleration library required for the task to the OpenCL standard; Using the API interface of the first OpenCL computing acceleration library to implement integrated calls; Implement StreamExecutor's support for OpenCL and use StreamExecutor's OpenCL backend to implement integrated calls to the second OpenCL computing acceleration library.

2. The method according to claim 1, characterized in that The OpenCL Device support for the TensorFlow device layer based on the OpenCL programming standard includes: The classes with different functions of OpenCL Device are implemented by inheritance, and the functions of obtaining devices and related contexts during the calculation process of various operators are implemented by using the classes with different functions.

3. The method according to claim 1, characterized in that The OpenCL kernel and OpenCL operators that implement the corresponding computing functions based on the OpenCL programming standard with reference to the computing functions of the preset kernel in TensorFlow include: Encapsulate the OpenCL kernel into an OpenCL operator and register the OpenCL operator in TensorFlow.

4. The method according to claim 1, wherein The OpenCLElement-wise operator library implemented based on the OpenCL programming standard includes: Implement the underlying OpenCL kernel one by one for the Element-Wise operators; The underlying OpenCL kernel is encapsulated to generate an OpenCL operator.

5. The method according to claim 1, wherein Adapting the computing acceleration library required by the task to the OpenCL standard includes: Implement the corresponding underlying OpenCL kernel of the computing acceleration library's operators, encapsulate and generate the corresponding OpenCL operators, and register them in TensorFlow.

6. The method according to claim 1, characterized in that Implementing StreamExecutor's support for OpenCL and utilizing StreamExecutor's OpenCL backend to implement integrated calls to the second OpenCL computing acceleration library includes: Implement the OpenCL Platform in StreamExecutor and register it with the TensorFlow framework; The OpenCL Platform is created according to the instructions and specifies the corresponding device Platform; Implement OpenCL Executor in StreamExecutor and register it with the TensorFlow framework; Complete the registration and access of the computing acceleration library in the LibrarySupport layer of StreamExecutor; In the TensorFlow framework, bind the corresponding operators in the computing acceleration library to the OpenCL device and register them.

7. The method according to claims 1-5, characterized in that The method further comprises: Optimize the implementation process of OpenCLKernel and OpenCL operators.

8. The method according to claim 7, characterized in that The optimization of the implementation process of the OpenCL kernel and the OpenCL operator includes: Use the singleton pattern to initialize the OpenCL platform, device, context, and command queue; Cache the compilation results of the OpenCL kernel program; Use reduction algorithms to optimize atomic operations.

9. An OpenCL TensorFlow implementation device, characterized in that: include: The first generation unit is used to implement OpenCL Device support of the TensorFlow device layer based on the OpenCL programming standard; The second generation unit is configured to implement an OpenCL kernel and an OpenCL operator corresponding to the computing function based on the OpenCL programming standard and with reference to the computing function of the preset kernel in TensorFlow; The third generation unit is used to implement the OpenCL Element-wise operator library based on the OpenCL programming standard; The adaptation unit is used to adapt the computing acceleration library required by the task to the OpenCL standard; A first calling unit is configured to implement an integrated call using an API interface of a first OpenCL computing acceleration library; The second calling unit is used to implement StreamExecutor's support for OpenCL, and utilize the OpenCL backend of StreamExecutor to implement integrated calling of the second OpenCL computing acceleration library.

10. The device according to claim 9, characterized in that The first generating unit is further configured to: The classes with different functions of OpenCL Device are implemented by inheritance, and the functions of obtaining devices and related contexts during the calculation process of various operators are implemented by using the classes with different functions.

11. The device according to claim 9, characterized in that The second generating unit is further configured to: Encapsulate the OpenCL kernel into an OpenCL operator and register the OpenCL operator in TensorFlow.

12. The device according to claim 9, characterized in that The third generating unit includes: A first generation module is used to implement the underlying OpenCL kernel of the Element-Wise operator one by one; The encapsulation module is used to encapsulate the device-side code and functions of the underlying OpenCL kernel to generate an OpenCL operator.

13. The device according to claim 9, characterized in that The adaption unit is further configured to: Implement the corresponding underlying OpenCL kernel of the computing acceleration library's operators, encapsulate and generate the corresponding OpenCL operators, and register them in TensorFlow.

14. The device according to claim 9, characterized in that The second calling unit includes: The second generation module is used to implement the OpenCL Platform in StreamExecutor and register it with the TensorFlow framework; A creation module is used to create the OpenCL Platform according to the instruction and specify the corresponding device Platform; The third generation module is used to implement OpenCL Executor in StreamExecutor and register it with the TensorFlow framework; The registration module is used to complete the registration and access of the computing acceleration library in the LibrarySupport layer of StreamExecutor; The binding module is used to bind and register the corresponding operators in the computing acceleration library with OpenCL devices in the TensorFlow framework.

15. The device according to claims 9-13, characterized in that The device further comprises: The optimization unit is used to optimize the implementation process of OpenCL kernels and OpenCL operators.

16. The device according to claim 15, characterized in that The optimization unit includes: Initialization module, used to implement the initialization of OpenCL platform, device, context and command queue OpenCL using singleton mode; Cache module, used to cache the compilation results of OpenCL kernel programs; The optimization module is used to optimize atomic operations using reduction algorithms.

17. An electronic device, characterized in that: include: at least one processor; as well as a memory communicatively connected to the at least one processor; wherein, The memory stores instructions that can be executed by the at least one processor, and the instructions are executed by the at least one processor to enable the at least one processor to perform the method according to any one of claims 1 to 8.

18. A non-transitory computer-readable storage medium storing computer instructions, characterized in that: The computer instructions are used to cause the computer to execute the method according to any one of claims 1-8.

19. A computer program product, characterized in that The invention comprises a computer program which, when executed by a processor, implements the method according to any one of claims 1 to 8.

Images