Method for converting C + + code into CUDA-C code

Abstract

The invention relates to the technical field of computers, and discloses a method for converting a C + + code into a CUDA-C code, and the method comprises the steps: receiving a C + + source code, carrying out lexical analysis and grammatical analysis, generating an abstract syntax tree, and carrying out semantic analysis to ensure that the source code is correct semantically; analyzing the abstract syntax tree, identifying a data dependency relationship among the calculation tasks, and determining loop statements and conditional statements in the source code; and converting a loop statement and a conditional statement in the source code into a kernel function, a thread block and a thread management code corresponding to the CUDA, and further generating a corresponding compilable source file. According to the method, the abstract syntax tree of the C + + code is analyzed, the parallelism in the abstract syntax tree is analyzed, the code is optimized, the CUDA-C efficient code is generated, then efficient automatic conversion of the code is achieved, and the method has great significance in accelerating program performance and shortening the development period.

Description

Technical Field

[0001] The present application relates to the field of computer technology, and in particular to a method for converting C++ code to CUDA-C code. Background Art

[0002] With the rapid development of computer technology, especially in the fields of big data processing, machine learning, scientific computing, etc., the requirements for computing performance are increasing day by day. The traditional CPU computing model seems to be unable to cope with large-scale data and high-complexity calculations, while GPU (graphics processing unit) has gradually become an important means to improve computing performance due to its powerful parallel processing and floating-point computing capabilities. CUDA (Compute Unified Device Architecture) is a parallel computing platform and programming model launched by NVIDIA, which allows developers to use NVIDIA's GPU for efficient parallel computing. CUDA-C is an extended C language based on the CUDA platform. It provides a wealth of library functions and programming interfaces to facilitate developers to write GPU-accelerated applications.

[0003] Although CUDA-C has significant advantages in parallel computing, many existing software systems and algorithms are written in C++. These C++ code libraries have been accumulated and optimized for a long time and have high reliability and performance advantages. If these C++ codes can be automatically converted to CUDA-C codes, it will be able to greatly expand the application scope of GPU acceleration technology and improve computing performance. However, there are significant differences between C++ and CUDA-C in syntax, memory management, parallel programming models, etc. Manually converting these codes is not only time-consuming and labor-intensive, but also prone to errors.

[0004] It should be noted that the information disclosed in the above background technology section is only used to enhance the understanding of the background of the present application, and therefore may include information that does not constitute the prior art known to ordinary technicians in the field. Summary of the invention

[0005] In order to provide a basic understanding of some aspects of the disclosed embodiments, a brief summary is given below. The summary is not an extensive review, nor is it intended to identify key / critical components or delineate the scope of protection of these embodiments, but rather serves as a prelude to the detailed description that follows.

[0006] The method for converting C++ code to CUDA-C code provided in the embodiment of the present disclosure includes:

[0007] Receive C++ source code;

[0008] Perform lexical analysis and syntax analysis on C++ source code, generate abstract syntax tree, and perform semantic analysis to ensure that the C++ source code is semantically correct;

[0009] Analyze the abstract syntax tree, identify data dependencies between various computing tasks, and determine loop statements and conditional statements in the C++ source code;

[0010] Convert loop statements and conditional statements in the C++ source code into kernel functions, thread blocks, and thread management codes corresponding to CUDA;

[0011] Generate corresponding compilable source files based on the converted CUDA corresponding kernel functions, thread blocks and thread management codes.

[0012] In some embodiments, the method further comprises:

[0013] Automatically handles memory allocation and access in CUDA, including the use of global memory, shared memory, and constant memory.

[0014] In some embodiments, generating corresponding compilable source files based on the converted CUDA corresponding kernel functions, thread blocks and thread management codes includes:

[0015] Reconstruct the corresponding compilable source file code generated by the converted CUDA kernel function, thread block and thread management code, optimize the memory access mode, reduce the number of global memory accesses, increase the use of shared memory and constant memory, optimize the configuration of thread blocks and threads according to the architectural characteristics of the GPU, and generate the corresponding compilable source file.

[0016] In some embodiments, after generating the corresponding compilable source file, the method further includes:

[0017] Use conditional compilation directives to provide a backwards compatible solution.

[0018] In some embodiments, after generating the corresponding compilable source file, the method further includes:

[0019] Based on the CUDA programming model architecture of the host CPU and the device GPU, the program corresponding to the generated compilable source file is run on the CPU. At this time, space is allocated in the host's memory to store data that needs to be directly processed by the CPU and data for ultimately receiving GPU processing results.

[0020] In some embodiments, the method further comprises:

[0021] Based on the program corresponding to the generated compilable source file, memory space is allocated on the GPU to store data and intermediate results processed on the GPU.

[0022] In some embodiments, the method further comprises:

[0023] Before starting parallel computing, copy the data prepared in the host memory to the GPU's device memory;

[0024] Start one or more kernel functions to execute on the GPU, where each kernel function processes a large number of data elements simultaneously to achieve data-level parallelism or task-level parallelism.

[0025] In some embodiments, the method further comprises:

[0026] When the GPU completes all calculations, the program corresponding to the generated compilable source file copies the result data in the device memory back to the host memory for further processing or output on the CPU;

[0027] The program corresponding to the generated compilable source file releases all previously allocated host and device memory after completing its work, where the allocated host and device memory includes data structures created on the host and memory space allocated on the device.

[0028] The method for converting C++ code to CUDA-C code provided in the embodiments of the present disclosure can achieve the following technical effects:

[0029] By analyzing the abstract syntax tree (AST) of the C++ code, analyzing the parallelism therein, generating a CUDA-C code string, and then optimizing the CUDA-C code in combination with the characteristics of the GPU hardware to generate efficient CUDA-C code. The automated conversion process greatly reduces the burden of developers manually writing CUDA-C code. Developers only need to focus on the logic and functions of the original C++ code without having to deeply understand the details of CUDA programming. At the same time, automated conversion helps reduce human errors and omissions and improve the quality and stability of the code. Ultimately, it improves development efficiency, accelerates computing performance, reduces development risks, promotes technology popularization, improves software performance, supports continuous optimization, and promotes cross-platform development, which brings significant convenience to developers and is of great significance for accelerating program performance and reducing development cycles.

[0030] The above general description and the following description are exemplary and explanatory only and are not intended to limit the present application. BRIEF DESCRIPTION OF THE DRAWINGS

[0031] One or more embodiments are exemplarily described by corresponding drawings, which do not limit the embodiments. Elements with the same reference numerals in the drawings are shown as similar elements, and the drawings do not constitute a scale limitation, and wherein:

[0032] Figure 1 It is a flowchart of a method for converting C++ code to CUDA-C code provided by an embodiment of the present disclosure;

[0033] Figure 2 It is a schematic diagram of a CUDA-C program running process provided by an embodiment of the present disclosure;

[0034] Figure 3 It is a flowchart of another method for converting C++ code to CUDA-C code provided by an embodiment of the present disclosure;

[0035] Figure 4 It is a flowchart of another method for converting C++ code to CUDA-C code provided by an embodiment of the present disclosure;

[0036] Figure 5 It is a structural diagram of a C++ code to CUDA-C code conversion system provided by an embodiment of the present disclosure. DETAILED DESCRIPTION

[0037] In order to be able to understand the features and technical contents of the embodiments of the present disclosure in more detail, the implementation of the embodiments of the present disclosure is described in detail below in conjunction with the accompanying drawings. The attached drawings are for reference only and are not used to limit the embodiments of the present disclosure. In the following technical description, for the convenience of explanation, a full understanding of the disclosed embodiments is provided through multiple details. However, one or more embodiments can still be implemented without these details. In other cases, to simplify the drawings, well-known structures and devices can be simplified for display.

[0038] The terms "first", "second", etc. in the embodiments of the present disclosure are used to distinguish similar objects, and are not necessarily used to describe a specific order or sequence. It should be understood that the terms used in this way can be interchanged where appropriate, so as to describe the embodiments of the present disclosure described herein. In addition, the terms "including" and "having" and any variations thereof are intended to cover non-exclusive inclusions.

[0039] Unless otherwise stated, the term "plurality" means two or more.

[0040] In the embodiment of the present disclosure, the character " / " indicates that the preceding and following objects are in an "or" relationship. For example, A / B indicates: A or B.

[0041] The term "and / or" is a description of the association relationship between objects, indicating that three relationships can exist. For example, A and / or B means: A or B, or, A and B.

[0042] The term "correspondence" may refer to an association relationship or a binding relationship. The correspondence between A and B means that there is an association relationship or a binding relationship between A and B.

[0043] The following describes the method for converting C++ code to CUDA-C code provided by the embodiments of the present disclosure in conjunction with the accompanying drawings.

[0044] Figure 1 It is a flowchart of a method for converting C++ code to CUDA-C code provided by an embodiment of the present disclosure.

[0045] like Figure 1 As shown, the conversion method from C++ code to CUDA-C code may include:

[0046] S01, receiving C++ source code;

[0047] S02, perform lexical analysis and syntax analysis on the C++ source code, generate an abstract syntax tree, and perform semantic analysis to ensure that the C++ source code is semantically correct;

[0048] S03, analyzing the abstract syntax tree, identifying the data dependency between various computing tasks, and determining loop statements and conditional statements in the C++ source code;

[0049] S04, converting loop statements and conditional statements in the C++ source code into kernel functions, thread blocks and thread management codes corresponding to CUDA;

[0050] S05, generating corresponding compilable source files based on the converted CUDA corresponding kernel functions, thread blocks and thread management codes.

[0051] The present invention can identify the parallel computing part in C++ code and automatically map it to the parallel programming model of CUDA-C, while handling key issues such as memory management and data synchronization to ensure that the converted code can run correctly and efficiently on the CUDA platform. The realization of this technology will provide strong support for the development of high-performance computing, big data analysis and other fields.

[0052] In some embodiments, Figure 1 The method may also include:

[0053] Automatically handles memory allocation and access in CUDA, including the use of global memory, shared memory, and constant memory.

[0054] In some embodiments, the above-mentioned generation of corresponding compilable source files based on the converted CUDA corresponding kernel functions, thread blocks and thread management codes may include:

[0055] Reconstruct the corresponding compilable source file code generated by the converted CUDA kernel function, thread block and thread management code, optimize the memory access mode, reduce the number of global memory accesses, increase the use of shared memory and constant memory, optimize the configuration of thread blocks and threads according to the architectural characteristics of the GPU, and generate the corresponding compilable source file.

[0056] In some embodiments, after generating the corresponding compilable source file, the method may further include:

[0057] Use conditional compilation directives to provide a backwards compatible solution.

[0058] In some embodiments, after generating the corresponding compilable source file, the method may further include:

[0059] Based on the CUDA programming model architecture of the host CPU and the device GPU, the program corresponding to the generated compilable source file is run on the CPU. At this time, space is allocated in the host's memory to store data that needs to be directly processed by the CPU and data for ultimately receiving GPU processing results.

[0060] In some embodiments, the above method may further include:

[0061] Based on the program corresponding to the generated compilable source file, memory space is allocated on the GPU to store data and intermediate results processed on the GPU.

[0062] In some embodiments, the above method may further include:

[0063] Before starting parallel computing, copy the data prepared in the host memory to the GPU's device memory;

[0064] Start one or more kernel functions to execute on the GPU, where each kernel function processes a large number of data elements simultaneously to achieve data-level parallelism or task-level parallelism.

[0065] In some embodiments, the above method may further include:

[0066] When the GPU completes all calculations, the program corresponding to the generated compilable source file copies the result data in the device memory back to the host memory for further processing or output on the CPU;

[0067] The program corresponding to the generated compilable source file releases all previously allocated host and device memory after completing its work, where the allocated host and device memory includes data structures created on the host and memory space allocated on the device.

[0068] Figure 2 is a schematic diagram of a CUDA-C program running process provided by an embodiment of the present disclosure, Figure 3 It is a flowchart of another method for converting C++ code to CUDA-C code provided by an embodiment of the present disclosure.

[0069] Combination Figure 2 and Figure 3 ,right Figure 1 In a specific example, the CUDA programming model uses a host and device architecture, where the host is usually the CPU and the device is the GPU. CUDA programs include code executed on the host and code executed on the device (called a "kernel" or "kernel"). Developers need to write kernel functions to define operations that are executed in parallel on the GPU. The process of running a CUDA-C program is as follows: Figure 2 As shown, the specific steps are as follows:

[0070] 1. Host memory allocation: The CUDA program first runs on the CPU (called the host), at which time space is allocated in the host's memory to store data that needs to be processed directly by the CPU and data that will eventually receive the GPU processing results.

[0071] 2. Device memory allocation: CUDA programs allocate memory space on the GPU (called the device). This part of memory is specifically used to store data and intermediate results that will be processed on the GPU. The memory on the GPU is usually faster than the CPU memory, but the capacity may be smaller and the access cost (such as latency and bandwidth) varies in different contexts.

[0072] 3. Host to device data copy: Before starting parallel computing, the prepared data (input data) in the host memory needs to be copied to the GPU's device memory. This step may introduce additional latency because it involves data transfer between different physical locations.

[0073] 4. Kernel call: Start one or more kernel functions to execute on the GPU. Each kernel function can process a large number of data elements at the same time, achieving data-level parallelism or task-level parallelism. The GPU uses its many computing cores (possibly thousands) to accelerate this process.

[0074] 5. Copy from device to host: When the GPU completes all calculations, the CUDA program needs to copy the result data in the device memory back to the host memory for further processing or output on the CPU.

[0075] 6. Memory release: To ensure efficient use of system resources, a CUDA program should release all previously allocated host and device memory after completing its work. This includes deleting data structures created on the host and memory space allocated on the device.

[0076] Based on the above CUDA programming model, the present invention performs lexical analysis and syntax analysis on the input C++ source code, generates an abstract syntax tree, identifies parallel code according to the abstract syntax tree, further analyzes the parallelism of the parallel code, and converts the C++ code into CUDA-C code by combining the syntax characteristics of CUDA-C and the CUDA programming model. The process of this method can be further described as follows Figure 3 As shown, the specific steps are as follows:

[0077] 1. Input the original source code, which is provided by the user for the compiler to use.

[0078] 2. Code parsing. The compiler front end performs lexical analysis and syntax analysis on the source code, generates an abstract syntax tree, and performs semantic analysis to ensure that the code is semantically correct, such as variables have been declared before use, types match, etc. This helps avoid errors during the conversion process.

[0079] 3. Parallelism analysis. By analyzing the abstract syntax tree, the data dependencies between various computing tasks are identified, which is the key to determining which code blocks can be executed in parallel. Focus on identifying loop structures that can be parallelized, such as nested loops, loops that can be iterated independently, etc. Consider optimization strategies such as loop unrolling and blocking to improve parallel performance.

[0080] 4. Code generation. Convert control structures, such as loops and conditional statements in C++, into CUDA-specific kernel functions, thread blocks, and thread management codes. Automatically handle memory allocation and access in CUDA, including the use of global memory, shared memory, and constant memory. Optimize memory access patterns and reduce memory latency.

[0081] 5. Code optimization. Refactor the generated CUDA-C code to reduce unnecessary calculations and data transfers and improve computing efficiency. Optimize memory access patterns, reduce the number of global memory accesses, and increase the use of shared memory and constant memory to reduce memory latency and improve bandwidth utilization. Rationally allocate register resources to avoid performance degradation caused by register overflow. According to the architectural characteristics of the GPU, optimize the configuration of thread blocks and threads to achieve better load balancing and higher resource utilization.

[0082] 6. Code output. Output the converted and optimized CUDA-C code as a compilable source file. Ensure that the generated CUDA-C code is compatible with the current CUDA version, and provide backward compatible solutions using conditional compilation instructions, etc.

[0083] The code conversion from C++ to CUDA-C can be completed through the above steps. The core of the present invention is to first analyze the source code and generate an abstract syntax tree, then analyze the abstract syntax tree to identify the parallelism, combine the syntax and programming model of CUDA-C, convert the identified parallel code into CUDA-C code, then reconstruct the CUDA-C code, optimize the configuration of thread blocks and threads, and finally efficiently convert it into CUDA-C code that can utilize the parallel computing capabilities of GPU.

[0084] The beneficial effects of the present invention are as follows: by analyzing the abstract syntax tree (AST) of the C++ code, analyzing the parallelism therein, generating a CUDA-C code string, and then optimizing the CUDA-C code in combination with the characteristics of the GPU hardware to generate CUDA-C efficient code. The automated conversion process greatly reduces the burden of developers manually writing CUDA-C code. Developers only need to focus on the logic and functions of the original C++ code without having to deeply understand the details of CUDA programming. At the same time, automated conversion helps to reduce human errors and omissions and improve the quality and stability of the code. Ultimately, it improves development efficiency, accelerates computing performance, reduces development risks, promotes technology popularization, improves software performance, supports continuous optimization, and promotes cross-platform development, bringing significant convenience to developers.

[0085] Figure 4 FIG. 1 is a flow chart of another method for converting C++ code to CUDA-C code provided by an embodiment of the present disclosure. Figure 4 As shown, in a specific example, taking the following C++ source code as an example, the source code Add.cpp input by the present disclosure is as follows:

[0086] void matrixAdd(const std::vector <float>&A,const std::vector <float>&B,std::vector <float>&C,int n){

[0087] for(int i=0; i <n;++i){

[0088] for(int j=0; j <n;++j){

[0089] int idx = i*n+j;

[0090] C[idx]=A[idx]+B[idx];

[0091] }

[0092] }

[0093] }

[0094] The above C++ code uses a for loop to perform the addition of two matrices of the same size, which is executed sequentially in the CPU, and the computing performance becomes a bottleneck. Even through compilation optimization and using vector instruction sets to improve performance, there is still a considerable performance gap because the parallel granularity of the CPU is coarser than that of the GPU and is limited by the number of cores and thread management capabilities.

[0095] The present invention is based on the secondary development of CLANG compiler, supports lexical analysis, syntax analysis, and semantic analysis of C++ code, and adds the code generation backend of CUDA-C. The running command is as follows:

[0096] Clang-emit-cuda Add.cpp-o Cuda-Add.cu–gencode arch=compute_61,code=sm_61

[0097] The -emit-cuda option specifies the generation of cuda code, -gencode arch specifies the virtual architecture of the gpu, and code specifies the physical architecture of the gpu.

[0098] According to the method provided by the present invention, code parsing is performed. First, lexical analysis is performed to scan the above source code and output a series of tags, including void, matrixAdd, (, const, std::vector <float>,&,A,,,B,C,int,n,{,for,int,i,=,0,;,etc. Perform syntax analysis based on the above tags and output an abstract syntax tree, including function declaration nodes, parameter list nodes, loop structure nodes (for loops), conditional judgment nodes (although these are actually boundary conditions of the loop), assignment expression nodes, etc. Perform semantic analysis based on the abstract syntax tree to check whether the syntax structure in the abstract syntax tree conforms to the semantic rules of the language, such as type matching, scope rules, etc. For the matrixAdd function, this step will confirm that A, B, and C are std::vector <float>type reference, n is of type int, and the type compatibility of the assignment operation.

[0099] Perform parallel analysis on the abstract syntax tree. There is a for loop node in the abstract syntax tree generated by the matrixAdd function. Perform parallel analysis on it and create a dependency graph. The circular dependency includes data dependency and control dependency. In the matrixAdd loop, the calculation of each element is independent and does not depend on the calculation results of other elements, so there is no data dependency. At the same time, the control flow of the loop is also fixed, there is no conditional branch or dependency between loop iterations, so there is no control dependency. Therefore, there is no circular dependency in the execution statement in the above for loop node.

[0100] According to the above parallel analysis, the for loop node in the abstract syntax tree is suitable for parallel computing. Add the CUDA-C backend in CLANG to generate a CUDA-C string. The specific CUDA-C string is as follows:

[0101] __global__void matrixAddKernel(const float*A,const float*B,float*C,int n){

[0102] int i=blockIdx.y*blockDim.y+threadIdx.y;

[0103]

[0104] The CUDA-C code generated here is not optimized for a specific architecture, so the CUDA-C backend of CLANG also has SUB-TARGET to optimize CUDA-C for the characteristics of each architecture. Because the compute capability of the hardware specified in the command line is 6.1, GPUs with a compute capability of 6.1 (such as Volta and Turing) are generally able to process larger thread blocks because they have higher parallel processing capabilities and larger resources. In NVIDIA GPUs, threads are scheduled in units of warps (32 threads per warp). Therefore, the size of the thread block is preferably a multiple of the warp size to avoid splitting the threads within the warp to different SMs (stream multiprocessors), which will increase the scheduling overhead. Therefore, the size of each thread block is adjusted to 16*16 here.

[0105] Finally, the CUDA-C code is output to the Cuda-Add.cu file.

[0106] It should be noted that, for the aforementioned method embodiments, for the sake of simplicity of description, they are all expressed as a series of action combinations, but those skilled in the art should be aware that the present disclosure is not limited by the order of the actions described, because according to the present disclosure, certain steps can be performed in other orders or simultaneously. Secondly, those skilled in the art should also be aware that the embodiments described in the specification are all optional embodiments, and the actions and modules involved are not necessarily required by the present disclosure.

[0107] Combination Figure 5 As shown, the embodiment of the present disclosure provides a C++ code to CUDA-C code conversion system 500, including a processor 504 and a memory 501. Optionally, the system may also include a communication interface 502 and a bus 503. Among them, the processor 504, the communication interface 502, and the memory 501 can communicate with each other through the bus 503. The communication interface 502 can be used for information transmission. The processor 504 can call the logic instructions in the memory 501 to execute the conversion method of C++ code to CUDA-C code in the above embodiment.

[0108] In addition, the logic instructions in the memory 501 described above can be implemented in the form of software functional units and can be stored in a computer-readable storage medium when sold or used as an independent product.

[0109] The memory 501 is a computer-readable storage medium that can be used to store software programs and computer executable programs, such as program instructions / modules corresponding to the method in the embodiment of the present disclosure. The processor 504 executes the function application and data processing by running the program instructions / modules stored in the memory 501, that is, the conversion method from C++ code to CUDA-C code in the above embodiment is implemented.

[0110] The memory 501 may include a program storage area and a data storage area, wherein the program storage area may store an operating system and an application required for at least one function; the data storage area may store data created according to the use of the terminal device, etc. In addition, the memory 501 may include a high-speed random access memory and may also include a non-volatile memory.

[0111] An embodiment of the present disclosure provides a computer-readable storage medium storing computer-executable instructions, wherein the computer-executable instructions are configured to execute a method for converting C++ codes to CUDA-C codes.

[0112] The computer-readable storage medium mentioned above may be a transient computer-readable storage medium or a non-transitory computer-readable storage medium.

[0113] The technical solution of the embodiment of the present disclosure can be embodied in the form of a software product, which is stored in a storage medium and includes one or more instructions for causing a computer device (which may be a personal computer, a server, or a network device, etc.) to execute all or part of the steps of the method of the embodiment of the present disclosure. The aforementioned storage medium may be a non-transient storage medium, including: a USB flash drive, a mobile hard disk, a read-only memory (ROM, Read-Only Memory), a random access memory (RAM, Random Access Memory), a disk or an optical disk, and other media that can store program codes, or a transient storage medium.

[0114] The above description and the accompanying drawings fully illustrate the embodiments of the present disclosure so that those skilled in the art can practice them. Other embodiments may include structural, logical, electrical, process and other changes. The embodiments represent possible variations only. Unless explicitly required, individual components and functions are optional, and the order of operations may vary. Parts and features of some embodiments may be included in or replace parts and features of other embodiments. As used in the description of the embodiments, unless the context clearly indicates otherwise, the singular forms "a", "an" and "the" are intended to include plural forms as well. Similarly, the term "and / or" as used in this application refers to any and all possible combinations of one or more associated listings. In addition, when used in this application, the term "comprise" and its variants "comprises" and / or including (comprising) refer to the existence of stated features, wholes, steps, operations, elements, and / or components, but do not exclude the existence or addition of one or more other features, wholes, steps, operations, elements, components and / or groups of these. In the absence of further restrictions, an element defined by the sentence "comprising a ..." does not exclude the existence of other identical elements in the process, method or device comprising the element. In this article, each embodiment may focus on the differences from other embodiments, and the same or similar parts between the embodiments may refer to each other. For the methods, products, etc. disclosed in the embodiments, if they correspond to the method part disclosed in the embodiments, then the relevant parts may refer to the description of the method part.

[0115] Those skilled in the art will appreciate that the units and algorithm steps of each example described in conjunction with the embodiments disclosed herein can be implemented in electronic hardware, or a combination of computer software and electronic hardware. Whether these functions are performed in hardware or software may depend on the specific application and design constraints of the technical solution. Technicians can use different methods to implement the described functions for each specific application, but such implementation should not be considered to be beyond the scope of the embodiments of the present disclosure. Technicians can clearly understand that for the convenience and simplicity of description, the specific working process of the systems, devices and units described above can refer to the corresponding process in the aforementioned method embodiment, and will not be repeated here.

[0116] In the embodiments disclosed herein, the disclosed methods and products (including but not limited to devices, equipment, etc.) can be implemented in other ways. For example, the device embodiments described above are only schematic. For example, the division of units can be only a logical function division. There may be other division methods in actual implementation, such as multiple units or components can be combined or integrated into another system, or some features can be ignored or not executed. In addition, the coupling or direct coupling or communication connection between each other shown or discussed can be through some interfaces, indirect coupling or communication connection of devices or units, which can be electrical, mechanical or other forms. The units described as separate components may or may not be physically separated, and the components displayed as units may or may not be physical units, that is, they may be located in one place, or they may be distributed on multiple network units. Some or all of the units may be selected according to actual needs to implement this embodiment. In addition, each functional unit in the embodiment of the present disclosure may be integrated in a processing unit, or each unit may exist physically alone, or two or more units may be integrated in one unit.

[0117] The flowchart and block diagram in the accompanying drawings show the possible architecture, functions and operations of the system, method and computer program product according to the embodiment of the present disclosure. In this regard, each box in the flowchart or block diagram can represent a module, a program segment or a part of the code, and a part of the module, program segment or code contains one or more executable instructions for realizing the specified logical function. In some alternative implementations, the functions marked in the box can also occur in a different order from the order marked in the accompanying drawings. For example, two consecutive boxes can actually be executed substantially in parallel, and they can sometimes be executed in the opposite order, which can depend on the functions involved. In the description corresponding to the flowchart and block diagram in the accompanying drawings, the operations or steps corresponding to different boxes can also occur in a different order from the order disclosed in the description, and sometimes there is no specific order between different operations or steps. For example, two consecutive operations or steps can actually be executed substantially in parallel, and they can sometimes be executed in the opposite order, which can depend on the functions involved. Each block in the block diagrams and / or flowcharts, and combinations of blocks in the block diagrams and / or flowcharts, may be implemented by a dedicated hardware-based system that performs the specified functions or actions, or may be implemented by a combination of dedicated hardware and computer instructions.< / float> < / float> < / float> < / float> < / float>

Claims

1. A method for converting C++ code to CUDA-C code, characterized in that: The method comprises: Receive C++ source code; Perform lexical analysis and syntax analysis on C++ source code, generate abstract syntax tree, and perform semantic analysis to ensure that the C++ source code is semantically correct; Analyze the abstract syntax tree, identify data dependencies between various computing tasks, and determine loop statements and conditional statements in the C++ source code; Convert loop statements and conditional statements in the C++ source code into kernel functions, thread blocks, and thread management codes corresponding to CUDA; Generate corresponding compilable source files based on the converted CUDA corresponding kernel functions, thread blocks and thread management codes.

2. The method according to claim 1, characterized in that The method further comprises: Automatically handles memory allocation and access in CUDA, including the use of global memory, shared memory, and constant memory.

3. The method according to claim 1, characterized in that The kernel function, thread block and thread management code corresponding to the converted CUDA generate corresponding compilable source files, including: Reconstruct the corresponding compilable source file code generated by the converted CUDA kernel function, thread block and thread management code, optimize the memory access mode, reduce the number of global memory accesses, increase the use of shared memory and constant memory, optimize the configuration of thread blocks and threads according to the architectural characteristics of the GPU, and generate the corresponding compilable source file.

4. The method according to claim 1, characterized in that: After generating the corresponding compilable source file, the method further includes: Use conditional compilation directives to provide a backwards compatible solution.

5. The method according to claim 1, characterized in that: After generating the corresponding compilable source file, the method further includes: Based on the CUDA programming model architecture of the host CPU and the device GPU, the program corresponding to the generated compilable source file is run on the CPU. At this time, space is allocated in the host's memory to store data that needs to be directly processed by the CPU and data for ultimately receiving GPU processing results.

6. The method according to claim 5, characterized in that The method further comprises: Based on the program corresponding to the generated compilable source file, memory space is allocated on the GPU to store data and intermediate results processed on the GPU.

7. The method according to claim 6, characterized in that The method further comprises: Before starting parallel computing, copy the data prepared in the host memory to the GPU's device memory; Start one or more kernel functions to execute on the GPU, where each kernel function processes a large number of data elements simultaneously to achieve data-level parallelism or task-level parallelism.

8. The method according to claim 7, characterized in that The method further comprises: When the GPU completes all calculations, the program corresponding to the generated compilable source file copies the result data in the device memory back to the host memory for further processing or output on the CPU; The program corresponding to the generated compilable source file releases all previously allocated host and device memory after completing its work, where the allocated host and device memory includes data structures created on the host and memory space allocated on the device.

Images