Offload Server, Offload Control Method, and Offload Program

Abstract

This offload server (1) comprises: a performance measurement unit (118) for compiling an application program of a parallel processing pattern, positioning the application program in a device for accelerator verification, and executing a process for performance measurement when offloading to an accelerator has occurred; and a positioning setting unit (170) that, when positioning a converted application program in one of a cloud server, a carrier edge server, and a user edge server on a network in accordance with a condition pertaining to a cost or a response time as designated by a user, calculates and sets the location for positioning of the application program on the basis of a linear planning formula in which device and link costs, a calculation resource upper limit, and a bandwidth upper limit are used as constraint conditions and the cost or response time of the calculation resource is used as an objective function.

Description

Technical Field

[0001] The present invention relates to an offloading server, an offloading control method, and an offloading program that automatically offload functional processing to accelerators such as GPUs (Graphics Processing Units) and FPGAs (Field Programmable Gate Arrays), and place the converted application program (hereinafter, appropriately referred to as an application) in an appropriate location.

Background Art

[0002] The use of heterogeneous computing resources other than CPUs (Central Processing Units) is increasing. For example, image processing is being performed on servers with enhanced GPUs (accelerators), and signal processing is beginning to be accelerated using FPGAs (accelerators). An FPGA is a programmable gate array whose configuration can be set by a designer or the like after manufacturing, and is a type of PLD (Programmable Logic Device). Amazon Web Services (AWS) (registered trademark) provides GPU instances and FPGA instances, and those resources can also be used on demand. Microsoft (registered trademark) is using FPGAs to improve the efficiency of searches.

[0003] The creation of diverse applications is expected using service cooperation technologies and the like, and higher performance of running applications can be expected by making better use of more advanced hardware. However, for this purpose, programming and settings according to the hardware to be operated are required. For example, a lot of technical knowledge such as CUDA (Compute Unified Device Architecture) and OpenCL (Open Computing Language) is required, and the hurdle is high. OpenCL is an open API (Application Programming Interface) that can uniformly handle all computing resources (not limited to CPUs and GPUs) without being tied to specific hardware.

[0004] In order to enable easy use of GPUs and FPGAs in a user's application, the following is required. That is, when deploying general-purpose applications such as image processing and cryptographic processing to be run in an environment, it is desired that the platform analyzes the application logic and automatically offloads the processing to GPUs and FPGAs.

[0005] The development environment CUDA for General Purpose GPUs (GPGPUs) that utilize the computing power of GPUs for purposes other than image processing is evolving. CUDA is a development environment for GPGPUs. Also, OpenCL has emerged as a standard specification for uniformly handling heterogeneous hardware such as GPUs, FPGAs, and many-core CPUs.

[0006] In CUDA and OpenCL, programming is performed through extensions of the C language. However, it is necessary to describe memory copying, freeing, etc. between devices such as GPUs and the CPU, and the difficulty of description is high. In fact, there are not many engineers who can proficiently use CUDA and OpenCL.

[0007] To easily perform GPGPU, there is a technology that is directive-based, specifies the locations for parallel processing such as loop statements, and the compiler converts it into device-specific code according to the directives. Examples of technical specifications include OpenACC (Open Accelerator), and examples of compilers include the PGI compiler (registered trademark). For example, in an example using OpenACC, the user specifies parallel processing, etc. in the code written in the C / C++ / Fortran language with OpenACC directives. The PGI compiler checks the parallelizability of the code, generates execution binaries for the GPU and the CPU, and modularizes the execution. The IBM JDK (registered trademark) supports the function of offloading parallel processing specified according to the lambda form of Java (registered trademark) to the GPU. By using these technologies, programmers do not need to be conscious of data allocation to the GPU memory, etc. In this way, technologies such as OpenCL, CUDA, and OpenACC enable offloading processing to GPUs and FPGAs.

[0008] However, even though offloading processing itself has become possible, there are many challenges in appropriate offloading. For example, there are compilers with an automatic parallelization function such as the Intel compiler (registered trademark). When performing automatic parallelization, parallel processing parts such as for loops (iteration statements) in the program are extracted. However, when operating in parallel using a GPU, due to the overhead caused by data transfer between the CPU-GPU memories, performance often cannot be achieved. When accelerating using a GPU, it is necessary for skilled personnel to tune with OpenCL or CUDA and search for appropriate parallel processing parts with a PGI compiler or the like. Therefore, it is difficult for users without skills to improve the performance of applications using a GPU, and even when using automatic parallelization technology, a lot of time is required until the start of use through trial-and-error tuning such as whether to parallelize a for loop or not.

[0009] Regarding placement, as an optimal utilization of network resources, there is research on optimizing the embedding position of VN (Virtual Network) for a group of servers on the network (see Non-Patent Document 1). In Non-Patent Document 1, the optimal placement of VN considering communication traffic is determined. However, it is targeted at virtual networks of single resources, with the aim of reducing the carrier's equipment cost and overall response time, and conditions such as the processing time of individual applications and the cost and response time requirements of individual users are not considered.

[0010] As an effort to automate the trial and error of parallel processing locations, Non-Patent Document 2 can be cited. Non-Patent Document 2 proposes environment-adaptive software that automatically performs conversion, resource configuration, etc. using the codes described once so that GPUs, FPGAs, many-core CPUs, etc. existing in the target environment can be utilized, with the aim of operating the application with high performance. Additionally, Non-Patent Document 2 proposes a method of automatically offloading the loop statements of application code to a GPU as an element of the environment-adaptive software and evaluates the performance improvement.

[0011] Non-Patent Document 3 proposes a method of automatically offloading the loop statements of application code to an FPGA as an element of the environment-adaptive software and evaluates the performance improvement. Non-Patent Document 4 evaluates a method of optimizing the resource amount (such as the number of virtual machine cores) for executing an application after automatically converting it for GPUs, etc. as an element of the environment-adaptive software.

Prior Art Documents

Non-Patent Documents

[0012]

Non-Patent Document 1

Non-Patent Document 2

[0013] In Non-Patent Documents 1 to 4, the evaluation focuses on shortening the processing time during automatic offloading. When offloading processing to heterogeneous devices such as GPUs and FPGAs, there is a problem that no proposal has been made regarding operating the converted application to meet user requirements (price, response time).

[0014] In view of such points, the present invention has been made, and when automatically converted so that it can be arranged on an offloading device such as a GPU or FPGA, the problem is to optimally arrange the converted application to meet the user's cost or response time requirements.

Means for Solving the Problems

[0015] To solve the above problems, an offloading server that offloads specific processing of an application program to an accelerator, comprising: an application code analysis unit that analyzes the source code of the application program; a data transfer specification unit that analyzes the reference relationship of variables used in the loop statements of the application program and performs data transfer specification using an explicit specification line that explicitly specifies data transfer outside the loop for data that can be transferred outside the loop; a parallel processing specification unit that identifies the loop statements of the application program and specifies and compiles parallel processing specification statements in the accelerator for each of the identified loop statements; a parallel processing pattern creation unit that creates a parallel processing pattern for specifying whether to perform parallel processing or not for loop statements that do not result in compilation errors while excluding loop statements that result in compilation errors from the offloading target; a performance measurement unit that compiles the application program of the parallel processing pattern and arranges it in an accelerator verification device and executes performance measurement processing when offloaded to the accelerator; and an arrangement setting unit that calculates and sets the arrangement location of the application program based on a linear programming formula with the costs of devices and links, the upper limit of computing resources, and the upper limit of bandwidth as constraint conditions and the cost of computing resources or response time as the objective function when arranging the converted application program on any of a cloud server, a carrier edge server, and a user edge server on the network according to the conditions of cost or response time specified by the user. The offloading server is characterized by comprising the above components.

Effect of the Invention

[0016] According to the present invention, when automatically converted so as to be deployable on an offloading device such as a GPU or FPGA, the converted application can be optimally arranged to meet the user's requirements for cost or response time.

Brief Description of the Drawings

[0017]

Figure 1

Figure 2

Figure 3

Figure 4

Figure 5

Figure 6

Figure 7

Figure 8

Figure 9A

Figure 9B

Figure 10

Figure 11

Figure 12

Figure 13

Figure 14

Figure 15

Figure 16

Figure 17

Figure 18

Embodiments for Carrying Out the Invention

[0018] Hereinafter, an offload server in a form for carrying out the present invention (hereinafter referred to as "the present embodiment") will be described with reference to the drawings. (Basic Concept of Automatic Offload of the Present Invention) In order to embody the concept of environment-adaptive software, the present inventor has hitherto proposed methods for GPU automatic offload of a loop statement of a program, FPGA automatic offload, and optimization of execution resources of a conversion application (see Non-Patent Documents 2, 3, and 4). Based on the study of the elemental technologies of these Non-Patent Documents 2, 3, and 4, the basic concept of the present invention will be described.

[0019] <Optimization of Resource Ratio between CPU and Offload Device> First, optimization of the resource ratio between the CPU and the offload device after the program conversion that can be offloaded to the device will be described. By the methods such as Non-Patent Document 2, an ordinary program can be automatically offloaded to an offload device such as a GPU or an FPGA. Currently, multi-core CPUs and many-core CPUs can flexibly allocate a certain percentage of all cores through virtualization using virtual machines or containers. Similarly, for GPUs, virtualization similar to that of CPUs has been carried out in recent years, and operations such as allocating a certain percentage of all cores of GPUs are becoming possible. Regarding FPGAs, the resource usage is often represented by the number of settings of Look Up Tables and Flip Flops, and unused gates can be used for other purposes.

[0020] In this way, it is possible to operate using a part of the total resources for CPUs, GPUs, and FPGAs. Appropriately optimizing the resources of CPUs and offload devices according to the application is important for improving cost performance.

[0021] Also, using the method of Non-Patent Document 2 etc., an application can be converted into code for CPU and GPU processing. However, even if the code itself is appropriate, if the resource amounts of the CPU and GPU are not in an appropriate balance, performance will not be achieved. For example, when performing a certain process, if the processing time of the CPU is 1000 seconds and the processing time of the GPU is 1 second, even if the process that can be offloaded is accelerated to a certain extent on the GPU, overall the CPU becomes the bottleneck.

[0022] Furthermore, in "K. Shirahata, H. Sato and S. Matsuoka, 'Hybrid Map Task Scheduling for GPU-Based Heterogeneous Clusters,' IEEE Second International Conference on Cloud Computing Technology and Science (CloudCom), pp.733 - 740, Dec. 2010." of Non-Patent Document 5, when processing tasks using the MapReduce (registered trademark) framework with a CPU and a GPU, the overall performance is improved by distributing Map tasks so that the execution times of the CPU and the GPU are the same.

[0023] The inventor of the present invention has conceived to determine the resource ratio between the CPU and the offload device as follows. That is, in order to avoid a bottleneck in processing on any device, referring to the above non-patent documents, etc., from the processing time of the test case, the resource ratio between the CPU and the offload device (hereinafter referred to as the "resource ratio") is determined so that the processing times of the CPU and the offload device are of the same order.

[0024] In addition, the inventor of the present invention adopts a method of gradually increasing the speed based on the performance measurement results in the verification environment during automatic offloading, like the method in Non-Patent Document 2. The reason is that regarding performance, it varies greatly not only depending on the code structure but also on the actual processing content such as the specifications of the hardware actually processed, the data size, and the number of loop iterations. Also, performance is difficult to predict statically and dynamic measurement is required. Therefore, when performing code conversion, since there are already performance measurement results in the verification environment, the resource ratio is determined using those results.

[0025] When performing performance measurement, measurement is carried out by specifying a test case. For example, when the processing time of the test case in the verification environment is CPU processing: 10 seconds and GPU processing: 5 seconds, the resources on the CPU side are considered to be twice as much for about the same processing time. Therefore, the resource ratio is 2:1. Regarding the user requirement of particularly wanting to speed up a certain process by offloading, a test case including that process is prepared, and the user requirement is reflected by speeding up the test case using the method in Non-Patent Document 2 or the like.

[0026] <Determination of the Resource Amounts of the CPU and the Offload Device and Automatic Verification> Next, the determination of the resource amounts (hereinafter referred to as the "resource amounts") of the CPU and the offload device and automatic verification will be described. When the resource ratio is determined by the above <Optimization of the Resource Ratio between the CPU and the Offload Device>, next, the application is deployed to the commercial environment. When deploying in a commercial environment, the resource amount is determined while keeping (maintaining) the resource ratio as much as possible to meet the cost requirements specified by the user. For example, regarding the CPU, assume that 1 VM costs 1000 yen per month, the GPU costs 4000 yen per month, and an appropriate resource ratio is 2:1. Also, assume that the user's budget is within 10000 yen per month. In this case, even with a resource ratio of 2:1, it fits within the user's budget of 10000 yen per month. So, the resource amount that keeps the appropriate resource ratio of 2:1, that is, the CPU is "2" and the GPU is "1", is secured and deployed in the commercial environment. Also, if the user's budget is within 5000 yen per month, the appropriate resource ratio of 2:1 cannot be maintained. In this case, as the resource amount, the CPU is "1" and the GPU is "1" are secured and deployed.

[0027] After securing resources and deploying the program in the commercial environment, in order to confirm that it operates before the user uses it, automatic verification is performed. In automatic verification, performance verification test cases and regression test cases are executed. Performance verification test cases are executed using a test automation tool such as Jenkins (registered trademark) for the assumed test cases specified by the user, and the processing time, throughput, etc. are measured. Regression test cases obtain information on software such as middleware and the OS installed in the system, and execute the corresponding regression tests using Jenkins or the like. Considerations for performing these automatic verifications with less preparation of test cases have been made in Non-Patent Document 6 (Y. Yamato, “Automatic verification technology of software patches for user virtual environments on IaaS cloud,” Journal of Cloud Computing, Springer, 2015, 4:4, DOI: 10.1186 / s13677-015-0028-6, Feb. 2015.) etc., and the technology of this Non-Patent Document 6 is used.

[0028] In the performance verification test case, it is checked whether the calculation results are incorrect even when offloading. Also, in the performance verification test case, the difference in calculation results between the case of not offloading and the case of offloading is also checked. For example, the PGI compiler that processes the GPU can confirm the difference in calculation results between the case of using the GPU and the case of not using the GPU through APIs such as PGI_compare (registered trademark) and acc_compare (registered trademark) of the PCAST (registered trademark) function. Note that even if parallel processing etc. is correctly offloaded, there may be cases where the calculation results do not completely match, such as different rounding errors between the GPU and the CPU. Therefore, for example, confirmation based on the IEEE 754 specification is performed, and the user is presented with whether the difference is acceptable and asked to confirm.

[0029] As a result of automatic verification, information such as the processing time and throughput of the performance verification test case, the calculation result difference, and the execution result of the regression test is presented to the user. The user is also presented with the resources secured (number and specifications of VMs, etc.) and their prices, and the user refers to this information to determine the start of operation.

[0030] <Resource, resource ratio, test case processing time> The resources, resource ratio, and test case processing time in this embodiment will be described. ·Regarding resources CPUs, GPUs, FPGAs, etc. have come to be provided as instances of virtual resources. As resources, there are the number of CPU cores, clock, memory capacity, disk size, the number of GPU cores, clock, memory capacity, and the gate scale of the FPGA (LE (registered trademark) in the case of Intel (registered trademark), LC (registered trademark) in the case of Xilinx (registered trademark) as the unit). Vendors such as cloud providers package them and provide them in the form of small-sized virtual machines or GPU instances. When virtualizing, the number of instances used can be said to be the amount of resources used.

[0031] · Regarding the resource ratio The ratio of the number of instances of CPU, GPU, and FPGA is the resource ratio. If the number of instances is 1, 2, or 3, the resource ratio is 1:2:3.

[0032] · Regarding the test case processing time This embodiment searches for and discovers an offloading pattern to speed up the test cases specified by the user. The test case is, for example, the number of transactions such as TPC-C (registered trademark) if it is a database, or the execution of Fourier transform processing on sample data if it is FFT. The processing time is the execution time when the sample processing is executed. For example, the processing time of process A was 10 seconds before offloading, but becomes 2 seconds after offloading. The execution times when executed on the CPU and when executed on the offloading device are respectively obtained in this form.

[0033] <Discovery of loop statements> Currently, it is difficult for the compiler to find that this loop statement is suitable for parallel processing on the GPU. It is difficult to predict the performance and power consumption by offloading to the GPU without actually measuring. Therefore, an instruction to offload this loop statement to the GPU is manually performed, and measurement trial and error is being carried out. The present invention automatically discovers an appropriate loop statement to offload to the GPU using a genetic algorithm (GA), which is an evolutionary computation method. That is, for a group of parallelizable loop statements, values are set to 1 when executed on the GPU and 0 when executed on the CPU for gene conversion, and repeated measurements are performed in the verification environment to search for an appropriate pattern.

[0034] (First Embodiment) Next, the offloading server 1, etc. in the form for implementing the present invention (hereinafter referred to as "this embodiment") will be described.

[0035] [Automatic GPU Offloading of Loop Statements] FIG. 1 is a functional block diagram showing a configuration example of the offload server 1 according to the first embodiment of the present invention. The offload server 1 is a device that automatically offloads specific processing of an application to an accelerator. As shown in FIG. 1, the offload server 1 includes a control unit 11, an input / output unit 12, a storage unit 13, and a verification machine 14 (Verification machine) (device for accelerator verification).

[0036] The input / output unit 12 includes a communication interface for transmitting and receiving information to and from each device, etc., and an input / output interface for transmitting and receiving information between input devices such as a touch panel and a keyboard, and output devices such as a monitor.

[0037] The storage unit 13 is composed of a hard disk, a flash memory, a RAM (Random Access Memory), etc., and stores a program (offload program) for executing each function of the control unit 11 and information (for example, an intermediate language file 133) necessary for the processing of the control unit 11 temporarily.

[0038] The storage unit 13 includes a test case DB (Test case database) 131, a facility resource DB 132, and an intermediate language file (Intermediate file) 133.

[0039] The test case DB 131 stores data of test items corresponding to the software to be verified. The data of the test items is, for example, data of a transaction test such as TPC-C in the case of a database system such as MySQL.

[0040] The equipment resource DB 132 holds resources such as servers held by the operator, pre-prepared information such as prices, and information on how much they are used. For example, there are 10 servers that can accommodate 3 GPU instances, 1 GPU instance costs 5000 yen per month, and out of the 10 servers, 2 servers, A and B, are fully utilized, and 1 server, C, has only 1 instance being used, etc. This information is used to determine the amount of resources to be secured when the user specifies operation conditions (conditions such as cost and performance). The user operation conditions are the cost conditions (e.g., a budget within 10000 yen per month, etc.) and performance conditions (e.g., what the transaction throughput such as TPC-C should be or within how many seconds for sample Fourier transform processing with 1 thread, etc.) specified by the user at the time of offloading request.

[0041] The intermediate language file 133 temporarily stores information necessary for the processing of the control unit 11 in the form of a programming language that intervenes between a high-level language and machine language.

[0042] The verification machine 14 is equipped with a CPU, a GPU, and an FPGA as a verification environment for the environment adaptation software.

[0043] The control unit 11 is an automatic offloading function unit (Automatic Offloading function) that controls the entire offloading server 1. The control unit 11 is realized, for example, by a CPU (Central Processing Unit), not shown, expanding and executing the application program (offloading program) stored in the storage unit 13 in the RAM.

[0044] The control unit 11 includes an application code specifying unit (Specify application code) 111, an application code analyzing unit (Analyze application code) 112, a data transfer specifying unit 113, a parallel processing specifying unit 114, a resource ratio determining unit 115, a resource amount setting unit 116, a placement setting unit 170, a parallel processing pattern creating unit 117, a performance measuring unit 118, an executable file creating unit 119, a production environment deployment unit (Deploy final binary files to production environment) 120, a performance measurement test extraction and execution unit (Extract performance test cases and run automatically) 121, and a user providing unit (Provide price and performance to a user to judge) 122.

[0045] <Application code specifying unit 111> The application code specifying unit 111 specifies the input application code. Specifically, the application code specifying unit 111 passes the application code described in the received file to the application code analyzing unit 112.

[0046] <Application code analyzing unit 112> The application code analyzing unit 112 analyzes the source code of the processing function and grasps the structure such as loop statements and FFT library calls.

[0047] <Data transfer specifying unit 113> The data transfer specifying unit 113 analyzes the reference relationships of variables used in the loop statements of the application program, and for data that can be transferred outside the loop, it performs data transfer specification using explicit specification lines (such as #pragma acc kernels, #pragma acc data copyin(a,b), #pragma acc data copyout(a,b), #prama acc parallel loop, #prama acc parallel loop vector to be described later) to explicitly specify data transfer outside the loop.

[0048] <Parallel processing specifying unit 114> The parallel processing specifying unit 114 identifies the loop statements (iteration statements) of the application program, and for each loop statement, it specifies a parallel processing specification statement in the accelerator and compiles it. The parallel processing specifying unit 114 includes an offload range extraction unit (Extract offload able area) 114a and an intermediate language file output unit (Output intermediate file) 114b.

[0049] The offload range extraction unit 114a identifies processes that can be offloaded to the GPU or FPGA, such as loop statements and FFT, and extracts intermediate language corresponding to the offload process. The intermediate language file output unit 114b outputs the extracted intermediate language file 133. The intermediate language extraction is not completed in one go, but is repeated for appropriate offload area exploration and for trial execution and optimization.

[0050] <Resource ratio determination unit 115> Based on the performance measurement results, the resource ratio determination unit 115 determines the processing times of the CPU and the offload device (test case CPU processing time and offload device processing time) as the resource ratio (described later). Specifically, the resource ratio determination unit 115 determines the resource ratio so that the processing times of the CPU and the offload device are of the same order. Also, when the difference between the processing times of the CPU and the offload device is equal to or greater than a predetermined threshold, the resource ratio determination unit 115 sets the resource ratio to a predetermined upper limit value.

[0051] <Resource amount setting unit 116> Based on the determined resource ratio, the resource amount setting unit 116 sets the resource amounts of the CPU and the offload device so as to satisfy a predetermined cost condition (described later). Specifically, the resource amount setting unit 116 sets the maximum resource amount that satisfies the predetermined cost condition while maintaining the determined resource ratio. Also, when the predetermined cost condition is not satisfied by setting the minimum resource amount while maintaining the determined resource ratio, the resource amount setting unit 116 breaks the resource ratio and sets the resource amounts of the CPU and the offload device to a smaller value (for example, the minimum) that satisfies the cost condition.

[0052] <Placement setting unit 170> When the configuration setting unit 170 deploys the converted application to any one of a cloud server, a carrier edge server, or a user edge server on the network according to the cost or response time conditions specified by the user, it takes the costs of devices and links, the upper limit of computing resources, and the upper limit of bandwidth as constraints, and calculates and sets the deployment location of the application based on a linear programming formula with the cost of computing resources or the response time as the objective function. Specifically, the configuration setting unit 170 calculates and sets the deployment destination (the deployment location of the APL) of the new application by a linear programming method based on the server in the facility resource DB 132, the specification information of the links, and the deployment information of the existing applications. In the linear programming method, for example, the objective function and constraints of the linear programming formula shown in [Equation 1] and [Equation 2] described later are used. The linear programming formulas shown in [Equation 1] and [Equation 2] described later are stored in the facility resource DB 132, read out by the configuration setting unit 170, and expanded on the memory processed by the configuration setting unit 170.

[0053] <Parallel processing pattern creation unit 117> The parallel processing pattern creation unit 117 creates a parallel processing pattern that excludes loop statements (iteration statements) that cause compilation errors from offloading targets and specifies whether to perform parallel processing or not for iteration statements that do not cause compilation errors.

[0054] <Performance measurement unit 118> The performance measurement unit 118 compiles the application program with the parallel processing pattern, deploys it to the verification machine 14, and executes the performance measurement process when offloaded to the accelerator. The performance measurement unit 118 includes a binary file deployment unit 118a. The binary file deployment unit 118a deploys (places) the executable file derived from the intermediate language to the verification machine 14 equipped with a GPU or FPGA.

[0055] The performance measurement unit 118 executes the arranged binary file, measures the performance when offloaded, and returns the performance measurement result to the offload range extraction unit 114a. In this case, the offload range extraction unit 114a performs extraction of another parallel processing pattern, and the intermediate language file output unit 114b attempts performance measurement based on the extracted intermediate language (see reference symbol a in FIG. 2 described later).

[0056] <Execution file creation unit 119> Based on the performance measurement results repeated a predetermined number of times, the execution file creation unit 119 selects a plurality of parallel processing patterns with high processing performance from a plurality of parallel processing patterns, creates another plurality of parallel processing patterns by crossing and mutating the parallel processing patterns with high processing performance. Then, the execution file creation unit 119 performs performance measurement up to newly, and after performance measurement a specified number of times, based on the performance measurement results, selects the parallel processing pattern with the highest processing performance from a plurality of parallel processing patterns, and compiles the parallel processing pattern with the highest processing performance to create an execution file.

[0057] <Production environment deployment unit 120> The production environment deployment unit 120 deploys the created execution file to the production environment for users ("deployment of the final binary file to the production environment"). The production environment deployment unit 120 determines the pattern that designates the final offload area and deploys it to the production environment for users.

[0058] <Performance measurement test extraction execution unit 121> After the execution file is deployed, the performance measurement test extraction execution unit 121 extracts performance test items from the test case DB 131 and executes the performance test ("deployment of the final binary file to the production environment"). After the execution file is deployed, the performance measurement test extraction execution unit 121 extracts performance test items from the test case DB 131 in order to show the performance to the user, and automatically executes the extracted performance test.

[0059] <User provision unit 122> The user providing unit 122 presents information such as price and performance to the user based on the performance test results (referred to as "providing information on price, performance, etc. to the user"). The test case DB 131 stores performance test items. The user providing unit 122 presents data such as price and performance to the user together with the above-mentioned performance test results based on the implementation results of the performance test corresponding to the test items stored in the test case DB 131. The user determines whether to start charging for the service based on the presented information such as price and performance. Here, for batch deployment to the production environment, the technology of Non-Patent Document 7 (Y. Yamato, M. Muroi, K. Tanaka and M. Uchimura, “Development of Template Management Technology for Easy Deployment of Virtual Resources on OpenStack,” Journal of Cloud Computing, Springer, 2014, 3:7, DOI: 10.1206 / s13677-014-0007-3, 12 pages, June 2014.) can be used, and for the automatic performance test, the technology of the aforementioned Non-Patent Document 6 can be used.

[0060] [Application of Genetic Algorithm] The offload server 1 can use GA (Genetic Algorithms) for offload optimization. The configuration of the offload server 1 when using GA is as follows. That is, the parallel processing specifying unit 114 sets the number of loop statements (repetition statements) that do not result in compilation errors as the gene length based on the genetic algorithm. The parallel processing pattern creating unit 117 maps the availability of accelerator processing to the gene pattern, where the case of performing accelerator processing is either 1 or 0, and the case of not performing it is the other 0 or 1.

[0061] The parallel processing pattern creation unit 117 prepares gene patterns of a specified number of individuals in which each value of the gene is randomly created as 1 or 0. The performance measurement unit 118 compiles the application code with the parallel processing specification statement in the accelerator according to each individual, and arranges it in the verification machine 14. The performance measurement unit 118 executes performance measurement processing in the verification machine 14.

[0062] Here, when the performance measurement unit 118 generates a gene of the same parallel processing pattern as before in an intermediate generation, it does not compile the application code corresponding to the parallel processing pattern and does not perform performance measurement, but uses the same value as the performance measurement value. In addition, for the application code in which a compilation error occurs and the application code for which the performance measurement does not end within a predetermined time, the performance measurement unit 118 sets the performance measurement value to a predetermined time (long time) as a timeout handling.

[0063] The execution file creation unit 119 performs performance measurement on all individuals, and evaluates so that the fitness becomes higher for individuals with shorter processing times. The execution file creation unit 119 selects, from all individuals, those with a fitness higher than a predetermined value (for example, the top n% of the total number, or the top m of the total number, where n and m are natural numbers) as high-performance individuals, and performs crossover and mutation processing on the selected individuals to create the next generation of individuals. After the processing of the specified number of generations is completed, the execution file creation unit 119 selects the parallel processing pattern with the highest performance as the solution.

[0064] Hereinafter, the automatic offloading operation of the offloading server 1 configured as described above will be described. [Automatic Offloading Operation] FIG. 2 is a diagram showing an automatic offloading process using the offloading server 1. As shown in FIG. 2, the offloading server 1 is applied to the elemental technologies of environmental adaptation software. The offloading server 1 includes a control unit (automatic offloading function unit) 11, a test case DB 131, a facility resource DB 132, an intermediate language file 133, and a verification machine 14. The offload server 1 acquires the application code 125 used by the user.

[0065] The user is, for example, a person who has contracted to use various devices (Device 151, device 152 having a CPU-GPU, device 153 having a CPU-FPGA, device 154 having a CPU). The offload server 1 automatically offloads the functional processing to the accelerators of the device 152 having a CPU-GPU and the device 153 having a CPU-FPGA.

[0066] Hereinafter, the operations of each part will be described with reference to the step numbers in FIG. 2. <Step S11: Specify application code> In step S11, the application code specifying unit 111 (see FIG. 1) passes the application code described in the received file to the application code analysis unit 112.

[0067] <Step S12: Analyze application code> In step S12, the application code analysis unit 112 (see FIG. 1) analyzes the source code of the processing function and grasps the structures such as loop statements and FFT library calls.

[0068] <Step S13: Extract offloadable area> In step S13, the parallel processing specifying unit 114 (see FIG. 1) specifies the loop statements (repetition statements) of the application, and for each repetition statement, specifies and compiles the parallel processing specification statements in the accelerator. Specifically, the offload range extraction unit 114a (see FIG. 1) specifies the processing that can be offloaded to the GPU or FPGA, such as loop statements and FFT, and extracts the intermediate language corresponding to the offload processing.

[0069] <Step S14: Output intermediate file> In step S14, the intermediate language file output unit 114b (see FIG. 1) outputs an intermediate language file 133. The intermediate language extraction is not completed in one go and is repeated for appropriate offload area search and for trial execution and optimization.

[0070] <Step S15: Compile error> In step S15, the parallel processing pattern creation unit 117 (see FIG. 1) creates a parallel processing pattern that excludes loop statements with compilation errors from offloading and specifies whether to perform parallel processing for iterative statements without compilation errors.

[0071] <Step S21: Deploy binary files> In step S21, the binary file placement unit 118a (see FIG. 1) deploys an executable file derived from the intermediate language to the verification machine 14 equipped with a GPU and an FPGA.

[0072] <Step S22: Measure performances> In step S22, the performance measurement unit 118 (see FIG. 1) executes the placed file and measures the performance when offloading. To make the offloading area more appropriate, this performance measurement result is returned to the offloading range extraction unit 114a, and the offloading range extraction unit 114a performs extraction of a different pattern. Then, the intermediate language file output unit 114b attempts performance measurement based on the extracted intermediate language (see reference symbol a in FIG. 2).

[0073] As shown by reference symbol a in FIG. 2, the control unit 11 repeatedly executes the above-described steps S12 to S22. The automatic offloading function of the control unit 11 is summarized as follows. That is, the parallel processing specification unit 114 identifies a loop statement (repetitive statement) of an application program, and for each repetitive statement, specifies a parallel processing specification statement on the GPU and compiles it. Then, the parallel processing pattern creation unit 117 excludes a loop statement that causes a compilation error from being an offloading target, and creates a parallel processing pattern for specifying whether to perform parallel processing or not for a loop statement that does not cause a compilation error. Then, the binary file placement unit 118a compiles the application program of the corresponding parallel processing pattern and places it in the verification machine 14, and the performance measurement unit 118 executes a performance measurement process on the verification machine 14. The execution file creation unit 119 selects a pattern with the highest processing performance from a plurality of parallel processing patterns based on the performance measurement results repeated a predetermined number of times, and compiles the selected pattern to create an execution file.

[0074] <Step S23: Resource amount setting according to user operation conditions> In step S23, the control unit 11 performs resource amount setting according to user operation conditions. That is, the resource ratio determination unit 115 of the control unit 11 determines the resource ratio between the CPU and the offloading device. Then, the resource amount setting unit 116 refers to the information in the equipment resource DB 132 based on the determined resource ratio, and sets the resource amounts of the CPU and the offloading device so as to satisfy the user operation conditions (to be described later with reference to FIG. 10).

[0075] <Step S24: Deploy final binary files to production environment> In step S24, the production environment placement unit 120 determines a pattern that designates the final offloading area and deploys it to the production environment for the user.

[0076] <Step S25: Extract performance test cases and run automatically> In step S25, after arranging the execution files, the performance measurement test extraction and execution unit 121 extracts performance test items from the test case DB 131 to show the performance to the user, and automatically executes the extracted performance tests.

[0077] <Step S26: Provide price and performance to a user to judge> In step S26, the user providing unit 122 presents information such as price and performance to the user based on the performance test results. The user judges whether to start charging for the service based on the presented information such as price and performance.

[0078] The above steps S11 to S26 are performed, for example, in the background of the user's service usage, and it is assumed that they are performed, for example, during the first day of trial use.

[0079] As described above, when the control unit (automatic offloading function unit) 11 of the offloading server 1 is applied to the elemental technologies of the environment adaptation software, in order to offload the functional processing, it extracts the offloading area from the source code of the application program used by the user and outputs the intermediate language (steps S11 to S15). The control unit 11 arranges and executes the execution file derived from the intermediate language on the verification machine 14 to verify the offloading effect (steps S21 to S22). After repeating the verification and determining the appropriate offloading area, the control unit 11 deploys the execution file to the actual production environment provided to the user and provides it as a service (steps S23 to S26).

[0080] [GPU Automatic Offloading Using GA] GPU automatic offloading is a process for repeatedly performing steps S12 to S22 in FIG. 2 on the GPU to finally obtain the offloading code to be deployed in step S23.

[0081] A GPU is a device that generally does not guarantee latency but is suitable for increasing throughput through parallel processing. Typical examples include encryption processing, image processing for camera video analysis, and machine learning processing for large-scale sensor data analysis, which involve a lot of repetitive processing. Therefore, the aim is to achieve high speed by automatically offloading the repetitive statements of an application to the GPU.

[0082] However, as described in the prior art, appropriate parallel processing is required for high speed. Especially when using a GPU, due to the memory transfer between the CPU and the GPU, performance often cannot be achieved unless the data size and the number of loop iterations are large. Also, depending on the timing of memory data transfer, etc., the combination of individual loop statements (repetitive statements) that can be parallelized and accelerated may not be the fastest. For example, in the case of 10 for statements (repetitive statements), when three of them, i.e., the 1st, 5th, and 10th, can be accelerated compared to the CPU, the combination of the 1st, 5th, and 10th may not necessarily be the fastest.

[0083] For appropriate parallel region specification, there have been attempts to optimize by trial and error regarding the parallelizability of for statements using the PGI compiler. However, trial and error takes a lot of operations, and when provided as a service, there are problems such as a delay in the start of user utilization and an increase in costs.

[0084] Therefore, in this embodiment, an appropriate offload region is automatically extracted from a general-purpose program that does not assume parallelization. For this purpose, first, a check for parallelizable for statements is performed, and then, for the group of parallelizable for statements, performance verification trials are repeated in a verification environment using GA to search for an appropriate region. By focusing on parallelizable for statements and holding and recombining the parallel processing patterns that can be accelerated in the form of gene parts, an efficient pattern that can be accelerated can be searched from the huge number of possible parallel processing patterns.

[0085] [Search image of the control unit (automatic offload function unit) 11 by Simple GA] Figure 3 is a diagram showing the search image of the control unit (automatic offloading function unit) 11 by Simple GA. Figure 3 shows the search image of the process and the gene sequence mapping of the for loop. GA is one of the combinatorial optimization methods that mimics the biological evolution process. The flowchart of GA is initialization → evaluation → selection → crossover → mutation → termination determination. In this embodiment, among GAs, Simple GA with simplified processing is used. In Simple GA, the genes are only 1 and 0, and roulette selection, one-point crossover, and mutation are to reverse the value of a single gene, which is a simplified GA.

[0086] <Initialization> In initialization, after checking the parallelizability of all for loops in the application code, the parallelizable for loops are mapped to the gene sequence. Set it to 1 if GPU processing is to be performed, and 0 if GPU processing is not to be performed. For the genes, prepare the specified number of individuals M and randomly assign 1 or 0 to each for loop. Specifically, the control unit (automatic offloading function unit) 11 (see Figure 1) acquires the application code 130 (see Figure 2) used by the user, and checks the parallelizability of the for loops from the code patterns 141 of the application code 130 as shown in Figure 3. As shown in Figure 3, if 5 for loops are found from the code patterns 141 (see reference numeral b in Figure 3), for each for loop, assign 1 digit, here 5 digits of 1 or 0 are randomly assigned to the 5 for loops. For example, set it to 0 when processing on the CPU and 1 when sending it to the GPU. However, at this stage, 1 or 0 is randomly assigned. The code corresponding to the gene length is 5 digits, and the code with a 5-digit gene length has 2 5 = 32 patterns, for example, 10001, 10010, …. In Figure 3, the circles (○ marks) in the code patterns 141 are shown as the image of the code.

[0087] <Evaluation> In the evaluation, deployment and performance measurement are performed (refer to symbol c in FIG. 3). That is, the performance measurement unit 118 (refer to FIG. 1) compiles the code corresponding to the gene, deploys it to the verification machine 14, and executes it. The performance measurement unit 118 performs benchmark performance measurement. The fitness of genes with good performance (parallel processing patterns) is increased.

[0088] <Selection> In the selection, high performance code patterns are selected based on the fitness (refer to symbol d in FIG. 3). The performance measurement unit 118 (refer to FIG. 1) selects genes with high fitness in a specified number of individuals based on the fitness. In this embodiment, roulette selection according to the fitness and elitist selection of the gene with the highest fitness are performed. In FIG. 3, the circles (○ marks) among the selected code patterns 142 are shown as a search image, indicating that they have been reduced to three.

[0089] <Crossover> In the crossover, at a certain crossover rate Pc, some genes are exchanged at a certain point among the selected individuals to create child individuals. The genes of a certain pattern (parallel processing pattern) selected by roulette selection are crossed with those of other patterns. The position of one-point crossover is arbitrary. For example, crossover is performed at the third digit among the above five-digit codes.

[0090] <Mutation> In the mutation, at a certain mutation rate Pm, each value of the gene of an individual is changed from 0 to 1 or from 1 to 0. Also, in order to avoid local solutions, mutations are introduced. Note that mutations may not be performed in order to reduce the amount of computation.

[0091] <End determination> As shown in Fig. 3, perform crossover and generation of the next-generation code pattern after mutation (refer to symbol e in Fig. 3). In the end determination, after repeating the specified number of generations T times, the process ends, and the gene with the highest fitness is taken as the solution. For example, measure the performance and select the three fastest, 10010, 01001, and 00101. Using these three with GA, the next generation performs recombination. For example, cross the first and the second to create a new pattern (parallel processing pattern) 11011. At this time, introduce mutations such as randomly changing 0 to 1 in the recombined pattern. Repeat the above to find the fastest pattern. Determine a specified number of generations (e.g., 20 generations), and take the pattern remaining in the final generation as the final solution.

[0092] <Deployment> Redeply in the production environment with the parallel processing pattern of the highest processing performance corresponding to the gene with the highest fitness and provide it to the user.

[0093] <Supplementary Explanation> Explain the case where there are a considerable number of for loops (loop statements; iteration statements) that cannot be offloaded to the GPU. For example, even if there are 200 for loops, only about 30 can be offloaded to the GPU. Here, exclude those that result in errors and perform GA on these 30.

[0094] OpenACC has a compiler that can extract GPU - oriented bytecode by specifying with the directive #pragma acc kernels, enabling GPU offloading during execution. By writing the for loop command in this #pragma, it is possible to determine whether the for loop can run on the GPU.

[0095] For example, when using C / C++, analyze the C / C++ code to find for-loops. Once a for-loop is found, write to it using #pragma acc kernels, #pragma acc parallel loop, or #pragma acc parallel loop vector, which are the grammar for parallel processing in OpenACC. Specifically, put each for-loop into #pragma acc kernels, #pragma acc parallel loop, or #pragma acc parallel loop vector and compile. If there is an error, it means that the for-loop cannot be processed by the GPU in the first place, so it is excluded.

[0096] In this way, find the remaining for-loops. Then, use those without errors as the length (gene length). If there are 5 for-loops without errors, the gene length is 5; if there are 10 for-loops without errors, the gene length is 10. Note that those that cannot be parallelized are cases where there is a data dependency such that the previous process is used for the next process. The above is the preparation stage. Next, perform GA processing.

[0097] Code patterns with gene lengths corresponding to the number of for-loops are obtained. Initially, randomly assign parallel processing patterns such as 10010, 01001, 00101, … GA processing is performed and compiled. At this time, even though a for-loop can be offloaded, an error may occur. This is the case when the for-loops are hierarchical (either one can be specified for GPU processing). In this case, the for-loop that results in an error may be left as it is. Specifically, there is a method of setting a timeout in a form that increases the processing time.

[0098] Deploy on the verification machine 14 and benchmark. For example, if it is image processing, benchmark with that image processing. The shorter the processing time, the higher the fitness is evaluated. For example, use the negative half-power of the processing time. If it takes 1 second for processing, the value is 1; if it takes 100 seconds, the value is 0.1; if it takes 0.01 seconds, the value is 10. Select those with high fitness. For example, select 3 to 5 out of 10, and recombine them to create a new code pattern. At this time, during the creation process, the same one as before may be created. In that case, since there is no need to perform the same benchmark, use the same data as before. In this embodiment, the code pattern and its processing time are stored in the storage unit 13. The search image of the control unit (automatic offloading function unit) 11 by the Simple GA has been described above. Next, a batch processing method for data transfer will be described.

[0099] [Batch Processing Method for Data Transfer] <Basic Concept> In addition to transferring the variables of the nested loop as high as possible to reduce CPU-GPU transfer, the present invention batches a large number of variable transfer timings and further reduces the transfers that the compiler automatically performs. In reducing transfers, not only at the nested unit level, but also for variables whose transfer timings to the GPU can be grouped, they are transferred in a batch. For example, for variables that are not used for processing the results of GPU processing on the CPU and then processing again on the GPU, it is also possible to transfer the variables defined on the CPU used in multiple loop statements to the GPU in a batch before the GPU processing starts and return them to the CPU after all GPU processing is completed.

[0100] In order to grasp the reference relationships of loops and variables during code analysis, for variables defined in multiple files as a result, for variables where GPU processing and CPU processing are not nested and can be separated, the specification to transfer them in a batch is specified using the OpenACC data copy statement. Variables that are transferred in a batch before the start of GPU processing and do not require transfer at the timing of loop statement processing are explicitly indicated as not requiring transfer using data present. When transferring data between the CPU and the GPU, a temporary area is created (#pragma acc declare create), the data is stored in the temporary area, and then the temporary area is synchronized (#pragma acc update) to instruct the transfer.

[0101] <Comparative Example> First, the comparative example will be described. The comparative examples are a normal CPU program (see Fig. 4), simple GPU utilization (see Fig. 5), and nested batching (Non-Patent Document 2) (see Fig. 6). Note that the <1> to <4>, etc. at the beginning of the loop statements in the following descriptions and figures are added for convenience of explanation (the same applies to other figures and their descriptions). The loop statement of the normal CPU program shown in Fig. 4 is described on the CPU program side. <1> Loop [for(i = 0; i < 10; i++)] { } Inside this, <2> Loop [for(j = 0; j < 20; j++ ] { exists. The symbol f in Fig. 4 is the setting of variables a and b in the above <2> loop. Also, <3> Loop [for(k = 0; k < 30; k++)] { } and <4> Loop [for(l = 0; l < 40; l++)] { } follow. The symbol g in Fig. 4 is the setting of variables c and d in the above <3> loop, and the symbol h in Fig. 4 is the setting of variables e and f in the above <4> loop. The normal CPU program shown in Fig. 4 is executed on the CPU (without using the GPU).

[0102] Fig. 5 is a diagram showing the loop statement when the normal CPU program shown in Fig. 4 is transferred from the CPU to the GPU using simple GPU utilization. The types of data transfer are data transfer from the CPU to the GPU and data transfer from the GPU to the CPU. Hereinafter, data transfer from the CPU to the GPU will be taken as an example. The loop statement of simple GPU utilization shown in Fig. 5 is described on the CPU program side. <1> Loop [for(i = 0; i < 10; i++)] { } Inside this, <2> There is a loop [for(j = 0; j < 20; j++)] { exists. Furthermore, as shown by the reference numeral i in FIG. 5, above the <1> loop [for(i = 0; i < 10; i++)] { }, a parallel - processable processing part such as a for statement by the PGI compiler is specified by the OpenACC directive #pragma acc kernels(parallel - process specification statement). As shown by the broken - line frame including the reference numeral i in FIG. 5, data is transferred from the CPU to the GPU by #pragma acc kernels. Here, since a and b are transferred at this timing, they are transferred 10 times.

[0103] Also, as shown by the reference numeral j in FIG. 5, above the <3> loop [for(k = 0; k < 30; k++)] { }, a parallel - processable processing part such as a for statement by the PGI compiler is specified by the OpenACC directive #pragma acc kernels. As shown by the broken - line frame including the reference numeral j in FIG. 5, c and d are transferred at this timing by #pragma acc kernels.

[0104] Here, above the <4> loop [for(l = 0; l < 40; l++)] { }, #pragma acc kernels is not specified. This loop is not processed by the GPU because the efficiency is poor even if it is processed by the GPU.

[0105] FIG. 6 is a diagram showing a loop statement in the case of data transfer from the CPU to the GPU and from the GPU to the CPU by nest - integration (Non - Patent Document 2). In the loop statement shown in FIG. 6, at the position indicated by the symbol k in FIG. 6, a data transfer instruction line from the CPU to the GPU, here the #pragma acc data copyin(a,b) of the copyin section of variables a and b, is inserted. Note that in this specification, for the sake of notation, parentheses () are attached to copyin(a,b). The same notation method will be adopted for the subsequent copyout(a,b) and datacopyin(a,b,c,d). The above #pragma acc data copyin(a,b) is specified at the topmost loop (here, the <1> loop [for(i=0; i<10; i++)]{ } above). Since a and b are transferred at the timing shown by the dashed line frame including the symbol k in FIG. 6, one transfer occurs.

[0106] Also, in the loop statement shown in FIG. 6, at the position indicated by the symbol l in FIG. 6, a data transfer instruction line from the GPU to the CPU, here the #pragma acc data copyout(a,b) of the copyout section of variables a and b, is inserted. The above #pragma acc data copyout(a,b) is specified below the <1> loop [for(i=0; i<10; i++)]{ }

[0107] In this way, in the data transfer from the CPU to the GPU, by inserting the #pragma acc data copyin(a,b) of the copyin section of variable a at the above-mentioned position, the data transfer is explicitly instructed. As a result, the data transfer can be performed in a batch at the highest possible loop, and it is possible to avoid the inefficient transfer of transferring data every time for each loop as in the loop statement of simple GPU utilization shown in FIG. 5.

[0108] <Embodiment> Next, this embodiment will be described. <Explicitly indicate variables that do not require transfer using data present> In this embodiment, for variables defined in multiple files, for variables where GPU processing and CPU processing are not nested and where CPU processing and GPU processing can be separated, a specification for batch transfer is specified using the data copy statement of OpenACC. At the same time, variables that are transferred in batch and for which transfer is not required at that timing are explicitly specified using data present.

[0109] FIG. 7 is a diagram showing a loop statement by batch transfer during CPU-GPU data transfer in this embodiment. FIG. 7 corresponds to the nested batching of FIG. 6 in the comparative example. In the loop statement shown in FIG. 7, at the position indicated by the reference numeral m in FIG. 7, a data transfer instruction line from the CPU to the GPU is inserted, here the #pragma acc datacopyin(a,b,c,d) of the copyin section of variables a, b, c, d. The above #pragma acc data copyin(a,b,c,d) is specified at the topmost loop (here, the <1> loop [for(i = 0; i < 10; i++)]{ }above) that does not include the setting and definition of variable a.

[0110] In this way, for variables defined in multiple files, for variables where GPU processing and CPU processing are not nested and where CPU processing and GPU processing can be separated, a specification for batch transfer is specified using the data copy statement #pragma acc data copyin(a,b,c,d) of OpenACC. At the timing shown by the dashed-dotted line frame including the reference numeral m in FIG. 7, a, b, c, d are transferred, so one transfer occurs.

[0111] And for variables that are transferred in batch using the above #pragma acc data copyin(a,b,c,d) and for which transfer is not required at that timing, the data present statement #pragma acc data present (a,b) is used to specify that variables already exist on the GPU at the timing shown by the double-dashed line frame including the reference numeral n in FIG. 7.

[0112] Using the above #pragma acc data copyin(a,b,c,d), they are transferred in a batch. For variables that do not require transfer at that timing, use the data present statement #pragma acc data present(c,d) to indicate that the variables already exist on the GPU at the timing shown by the two-dot chain line box including the symbol o in Figure 7. At the timing when the loops of <1> and <3> are processed on the GPU and the GPU processing is completed, insert the data transfer instruction line from the GPU to the CPU, here the #pragma acc data copyout(a,b,c,d) of the copyout section of variables a, b, c, and d, at the position p where the <3> loop in Figure 7 ends.

[0113] By specifying batch transfer for variables that can be transferred in a batch and explicitly indicating variables that have already been transferred and do not require transfer using data present, transfer can be reduced, and further efficiency improvement of the offloading means can be achieved. However, even if transfer is instructed with OpenACC, depending on the compiler, the compiler may automatically transfer data. Automatic transfer by the compiler refers to an event where, unlike the OpenACC instruction, the data is automatically transferred depending on the compiler even though transfer between the CPU and GPU is not originally required.

[0114] 《Temporary Area Storage of Data》 Figure 8 is a diagram showing a loop statement for batch transfer during CPU-GPU data transfer in this embodiment. Figure 8 corresponds to the nested batching and explicit indication of variables that do not require transfer in Figure 7. In the loop statement shown in Figure 8, at the position indicated by the symbol q in Figure 8, specify the #pragma acc declare create statement of OpenACC that creates a temporary area during CPU-GPU data transfer. Thereby, during CPU-GPU data transfer, a temporary area is created (#pragma acc declare create), and the data is stored in the temporary area.

[0115] Also, at the position indicated by the symbol r in FIG. 8, the transfer is instructed by specifying the OpenACC declare create statement #pragma acc update for synchronizing the temporary area.

[0116] In this way, by creating a temporary area, initializing parameters in the temporary area, and using it for CPU-GPU transfer, unnecessary CPU-GPU transfers can be blocked. Although not intended by the OpenACC instruction, transfers that degrade performance can be reduced.

[0117] [GPU Offloading Processing] By the above-described batch processing method for data transfer, appropriate loop statements for offloading can be extracted, and inefficient data transfer can be avoided. However, even when using the above-described batch processing method for data transfer, there are still programs that are not suitable for GPU offloading. For effective GPU offloading, it is necessary that the number of loop iterations of the processing to be offloaded is large.

[0118] Therefore, in this embodiment, as a preliminary stage of full-scale offloading processing exploration, the number of loop iterations is investigated using a profiling tool. When using a profiling tool, the number of executions of each line can be investigated, so that, for example, programs with loops of 50 million times or more can be pre-assigned as targets for offloading processing exploration. This will be specifically described below (partially overlapping with the content described in FIG. 2).

[0119] In this embodiment, first, the application code analysis unit 112 (FIG. 1) analyzes the application and grasps loop statements such as for, do, and while. Next, sample processing is executed, and using a profiling tool, the number of loop iterations of each loop statement is investigated, and a determination is made as to whether to conduct full-scale exploration based on whether there is a loop with a certain value or more.

[0120] When it is determined to conduct a full-scale exploration, enter the GA process (see Figure 2). In the initialization step, after checking the parallelizability of all loop statements in the application code, map the parallelizable loop statements to the gene sequence as 1 if they are to be processed by the GPU, and 0 if not. Genes are prepared for the specified number of individuals, and each value of the genes is randomly assigned 1 or 0.

[0121] Here, in the code corresponding to the gene, from the variable data reference relationship within the loop statement specified for GPU processing, add explicit instructions for data transfer (#pragma acc data copyin / copyout / copy).

[0122] In the evaluation step, compile the code corresponding to the gene, deploy it to the verification machine and execute it, and perform benchmark performance measurement. Then, increase the fitness of the genes with good performance patterns. The code corresponding to the gene has parallel processing instruction lines (for example, see reference f in Figure 4) and data transfer instruction lines (for example, see reference h in Figure 4, reference i in Figure 5, reference k in Figure 6) inserted as described above.

[0123] In the selection step, based on the fitness, select genes with high fitness for the specified number of individuals. In this embodiment, roulette selection according to the fitness and elite selection of the gene with the highest fitness are performed. In the crossover step, at a certain crossover rate Pc, exchange some genes at a certain point between the selected individuals to create offspring individuals. In the mutation step, at a certain mutation rate Pm, change each value of the genes of the individual from 0 to 1 or from 1 to 0.

[0124] When the mutation step is completed and the specified number of genes for the next generation is created, add explicit instructions for data transfer in the same way as in the initialization step, and repeat the evaluation, selection, crossover, and mutation steps.

[0125] Finally, in the end determination step, after performing a specified number of generations of iterations, the process is terminated, and the gene with the highest fitness is taken as the solution. The code pattern with the highest performance corresponding to the gene with the highest fitness is redeployed in the production environment and provided to the user.

[0126] Hereinafter, the implementation of the offloading server 1 will be described. This implementation is for verifying the effectiveness of the present embodiment. [Implementation] An implementation of automatically offloading a C / C++ application using a general-purpose PGI compiler will be described. In this implementation, since the purpose is to verify the effectiveness of GPU automatic offloading, the target application is an application in the C / C++ language, and for GPU processing itself, a conventional PGI compiler is used for the description.

[0127] The C / C++ language is popular in the development of OSS (Open Source Software) and proprietary software, and many applications are developed in the C / C++ language. To verify the offloading of applications used by general users, general-purpose applications of OSS such as cryptographic processing and image processing are utilized.

[0128] GPU processing is performed by the PGI compiler. The PGI compiler is a compiler for C / C++ / Fortran that interprets OpenACC. In the present embodiment, parallelizable processing parts such as for loops are specified with the OpenACC directive #pragma acc kernels (parallel processing specification statement). Thereby, GPU-oriented bytecode is extracted, and GPU offloading is made possible by its execution. Further, when there is a dependency between data within a for loop and the processing cannot be parallelized, or when different multiple levels of nested for loops are specified, an error is issued. Also, explicit data transfer instructions are made possible by directives such as #pragma acc data copyin / copyout / copy.

[0129] In accordance with the specification in the above #pragma acc kernels (parallel processing specification statement), an explicit data transfer instruction is given by inserting #pragma acc data copyout(a[…]) in the copyin section of OpenACC to the above-mentioned position.

[0130] <Overview of the implementation operation> The overview of the implementation operation will be described. The implementation performs the following processes. Before starting the process of the flow in FIGS. 9A - B below, prepare a C / C++ application to be accelerated and a benchmark tool for measuring its performance.

[0131] In the implementation, when there is a request to use a C / C++ application, first, the code of the C / C++ application is analyzed to discover for loops and to grasp the program structure such as variable data used within the for loops. For syntax analysis, a syntax analysis library such as LLVM / Clang is used.

[0132] In the implementation, first, to obtain an expectation as to whether the application has a GPU offloading effect, a benchmark is executed to grasp the number of loop iterations of the for loop grasped by the above syntax analysis. To grasp the number of loop iterations, gcov of GNU coverage or the like is used. As profiling tools, "GNU profiler (gprof)" and "GNU coverage (gcov)" are known. Since both can investigate the number of executions of each line, either can be used. The number of executions can be set, for example, to target only applications with a loop count of 10 million or more, but this value can be changed.

[0133] General-purpose applications for CPUs are not implemented assuming parallelization. Therefore, first, it is necessary to exclude for loops for which GPU processing itself is impossible. So, for each for loop one by one, try inserting the #pragma acc kernels, #pragma acc parallel loop, or #pragma acc parallel loop vector directives for GPU processing and determine whether an error occurs during compilation. There are several types of compilation errors. For example, when an external routine is called within a for loop, when different levels are redundantly specified in nested for loops, when there is a process to exit the for loop halfway through using break, etc., or when there is data dependency in the data of the for loop. Depending on the application, the types of compilation errors are diverse, and there are other cases as well. Compilation errors are excluded from the processing targets, and #pragma directives are not inserted.

[0134] Compilation errors are difficult to automatically handle, and often no effect is achieved even if handled. In the case of external routine calls, it may be possible to avoid them using #pragma acc routine, but many external calls are libraries, and even if GPU processing is included, that call becomes a bottleneck and no performance is achieved. Since each for loop is tried one by one, no compilation error occurs regarding nested errors. Also, when exiting halfway through using break, etc., it is necessary to fix the number of loop iterations for parallel processing, and program modification is required. If there is data dependency, parallel processing itself cannot be done in the first place.

[0135] Here, if the number of loop statements that do not result in an error even when parallel processed is a, then a becomes the gene length. Map the application code to a gene of length a, where 1 in the gene corresponds to having a parallel processing directive and 0 corresponds to not having it.

[0136] Next, as initial values, gene sequences of a specified number of individuals are prepared. Each value of the gene is created by randomly assigning 0 and 1 as described in Figure 3. According to the prepared gene sequence, when the value of the gene is 1, directives for specifying GPU processing, namely #pragma acc kernels, #pragma acc parallel loop, and #pragma acc parallel loop vector, are inserted into the C / C++ code. The reason for not making single loop, etc. parallel is that for the same processing, kernels have better performance as far as the PGI compiler is concerned. At this stage, within the code corresponding to a certain gene, the part to be processed by the GPU is determined.

[0137] The C / C++ code with directives for parallel processing and data transfer inserted is compiled using the PGI compiler on a machine equipped with a GPU. The compiled executable file is deployed, and the performance and power consumption are measured using a benchmark tool.

[0138] For all individuals, after benchmark performance measurement, the fitness of each gene sequence is set according to the benchmark processing time and power consumption. Based on the set fitness, the selection of individuals to be retained is performed. For the selected individuals, genetic algorithm (GA) processing such as crossover processing, mutation processing, and direct copy processing is carried out to create the next-generation population of individuals.

[0139] For the next-generation individuals, directive insertion, compilation, performance measurement, fitness setting, selection, crossover, and mutation processing are performed. Here, during the GA processing, if a gene of the same pattern as before occurs, for that individual, compilation and performance measurement are not performed, and the same measurement values as before are used.

[0140] After the GA processing for the specified number of generations is completed, the C / C++ code with directives corresponding to the gene sequence with the highest performance is taken as the solution.

[0141] Among these, the population size, number of generations, crossover rate, mutation rate, fitness setting, and selection method are parameters of the GA and are specified separately. The proposed technique enables the automation of GPU offloading, which conventionally required the time and skills of a specialist, by automating the above processing.

[0142] Figures 9A - B are flowcharts for explaining the operation overview of the above-described implementation, and Figures 9A and 9B are connected by connectors. Perform the following processing using an OpenACC compiler for C / C++.

[0143] <Code analysis> In step S101, the application code analysis unit 112 (see Figure 1) performs code analysis of the C / C++ application.

[0144] <Loop statement identification> In step S102, the parallel processing specification unit 114 (see Figure 1) identifies the loop statements and reference relationships of the C / C++ application.

[0145] <Parallel processing feasibility of loop statements> In step S103, the parallel processing specification unit 114 checks the GPU processing feasibility of each loop statement (#pragma acc kernels).

[0146] <Loop statement iteration> The control unit (automatic offloading function unit) 11 repeats the processing of steps S105 - S116 for the number of loop statements between the loop start of step S104 and the loop end of step S117.

[0147] <Iteration of the number of loops (Part 1)> The control unit (automatic offloading function unit) 11 repeats the processing of steps S106 - S107 for the number of loop statements between the loop start of step S105 and the loop end of step S108. In step S106, the parallel processing specification unit 114 specifies GPU processing (#pragma acc kernels) for each loop statement and compiles it with OpenACC. In step S107, when there is an error, the parallel processing specification unit 114 checks the GPU processing possibility with the next directive sentence (#pragma acc parallel loop).

[0148] <Loop number repetition (Part 2)> The control unit (automatic offloading function unit) 11 repeats the processing of steps S110 - S111 for the number of loop statements between the loop start of step S109 and the loop end of step S112. In step S110, the parallel processing specification unit 114 specifies GPU processing (#pragma acc parallel loop) for each loop statement and compiles it with OpenACC. In step S111, when there is an error, the parallel processing specification unit 114 checks the GPU processing possibility with the next directive sentence (#pragma acc parallel loop vector).

[0149] <Loop number repetition (Part 3)> The control unit (automatic offloading function unit) 11 repeats the processing of steps S114 - S115 for the number of loop statements between the loop start of step S113 and the loop end of step S116. In step S114, the parallel processing specification unit 114 specifies GPU processing (#pragma acc parallel loop vector) for each loop statement and compiles it with OpenACC. In step S115, when there is an error, the parallel processing specification unit 114 removes the GPU processing directive sentence from the loop statement.

[0150] <for statement number count> In step S118, the parallel processing specification unit 114 counts the number of loop statements (here, for statements) without compilation errors and uses it as the gene length.

[0151] <Specified individual number pattern preparation> Next, as an initial value, the parallel processing specifying unit 114 prepares gene sequences of a specified number of individuals. Here, they are created by randomly assigning 0 and 1. In step S119, the parallel processing specifying unit 114 maps the C / C++ application code to genes and prepares a specified number of individual patterns. According to the prepared gene sequences, when the value of a gene is 1, a directive for specifying parallel processing is inserted into the C / C++ code (for example, refer to the #pragma directive in FIG. 3).

[0152] The control unit (automatic offloading function unit) 11 repeats the processing of steps S121 - S130 for a specified number of generations between the loop start of step S120 and the loop end of step S131 in FIG. 9B. Also, in the repetition of the specified number of generations, the processing of steps S122 - S125 is repeated for a specified number of individuals between the loop start of step S121 and the loop end of step S126. That is, within the repetition of the specified number of generations, the repetition of the specified number of individuals is processed in a nested state.

[0153] <Data transfer specification> In step S122, the data transfer specifying unit 113 performs data transfer specification using explicit instruction lines (#pragma acc data copy / copyin / copyout / present and #pragam acc declarecreate, #pragma acc update) based on the variable reference relationship.

[0154] <Compilation> In step S123, the parallel processing pattern creation unit 117 (refer to FIG. 1) compiles the C / C++ code with directives specified according to the gene pattern using the PGI compiler. That is, the parallel processing pattern creation unit 117 compiles the created C / C++ code using the PGI compiler on the verification machine 14 equipped with a GPU. Here, a compilation error may occur, for example, when specifying multiple nested for loops in parallel. In this case, it is treated in the same way as when the processing time during performance measurement times out.

[0155] In step S124, the performance measurement unit 118 (see FIG. 1) deploys the executable file to the verification machine 14 equipped with a CPU-GPU. In step S125, the performance measurement unit 118 executes the deployed binary file and measures the benchmark performance when offloaded.

[0156] Here, in the intermediate generations, genes with the same pattern as before are not measured and the same values are used. That is, during the GA process, if a gene with the same pattern as before appears, for that individual, compilation and performance measurement are not performed, and the same measurement value as before is used.

[0157] In step S127, the performance measurement unit 118 (see FIG. 1) measures the processing time.

[0158] In step S128, the performance measurement unit 118 sets an evaluation value based on the measured processing time.

[0159] In step S129, the executable file creation unit 119 (see FIG. 1) evaluates so that the fitness is higher for individuals with shorter processing times and selects individuals with high performance. The executable file creation unit 119 selects, as a solution, a pattern with a short time and low power consumption amount among the measured multiple patterns.

[0160] In step S130, the executable file creation unit 119 performs crossover and mutation processes on the selected individuals to create the next generation of individuals. The executable file creation unit 119 performs compilation, performance measurement, fitness setting, selection, crossover, and mutation processes on the next generation of individuals. That is, for all individuals, after benchmark performance measurement, the fitness of each gene sequence is set according to the benchmark processing time. Selection of the remaining individuals is performed according to the set fitness. The executable file creation unit 119 performs GA processes of crossover processing, mutation processing, and direct copy processing on the selected individuals to create the next generation of individual groups.

[0161] In step S132, after the GA process for the specified number of generations ends, the execution file creation unit 119 uses the C / C++ code (the highest-performance parallel processing pattern) corresponding to the gene sequence with the highest performance as the solution.

[0162] <GA parameters> The above population size, number of generations, crossover rate, mutation rate, fitness setting, and selection method are GA parameters. The GA parameters may be set, for example, as follows. The parameters and conditions of the Simple GA to be executed can be, for example, as follows. Gene length: Number of parallelizable loop statements Population size M: Less than or equal to the gene length Number of generations T: Less than or equal to the gene length Fitness: (Processing time) (-1 / 2)

[0163] With this setting, the shorter the benchmark processing time, the higher the fitness. Also, by setting the fitness in a form that includes the (-1 / 2) power of the processing time, it is possible to prevent the fitness of a specific individual with a short processing time from becoming too high and the search range from becoming narrow. Also, if the performance measurement does not end within a certain time, a timeout is set, and the fitness is calculated assuming a time (a long time) such as 1000 seconds for the processing time. This timeout time may be changed according to the performance measurement characteristics. Selection: Roulette selection However, elitist preservation is also performed, in which the gene with the highest fitness in a generation is saved in the next generation without crossover or mutation. Crossover rate Pc: 0.9 Mutation rate Pm: 0.05

[0164] <Cost performance> The cost performance of the automatic offloading function will be described. Looking only at the hardware price of GPU boards such as NVIDIA Tesla, the price of a machine equipped with a GPU is about twice that of a machine with only a normal CPU. However, generally in terms of the costs of a data center etc., the costs of hardware and system development are less than 1 / 3, the operation costs such as electricity costs and maintenance / operation systems exceed 1 / 3, and other costs such as service orders are about 1 / 3. In this embodiment, for applications that perform time-consuming processes such as cryptographic processing and image processing, the processing performance can be improved by more than twice. Therefore, even if the server hardware price itself doubles, sufficient cost-effectiveness can be expected.

[0165] In this embodiment, gcov, gprof, etc. are used to identify in advance applications with many loops and long execution times, and offloading trials are conducted. Thereby, applications that can be efficiently accelerated can be found.

[0166] <Time until the start of production service use> The time until the start of production service use will be described. Assuming that it takes about 3 minutes for one performance measurement from compilation, in a GA with 20 individuals and 20 generations, it takes up to about 20 hours to search for solutions. However, since compilation and measurement of the same gene pattern as before are omitted, it will be completed within 8 hours. In fact, in many cloud, hosting, and network services, it takes about half a day to start using the service. In this embodiment, for example, automatic offloading within half a day is possible. Therefore, if it is automatic offloading within half a day, if trial use is possible at first, it can be expected to sufficiently improve user satisfaction.

[0167] In order to search the offload part in a shorter time, it is conceivable to measure the performance in parallel for the number of individuals using a plurality of verification machines. Depending on the application, adjusting the timeout time also leads to shortening the time. For example, when the offload process takes twice as long as the execution time on the CPU, it is set as the timeout. Also, the higher the number of individuals and the number of generations, the higher the possibility of finding a high-performance solution. However, when maximizing each parameter, it is necessary to perform compilation and performance benchmarking for the number of individuals × the number of generations. Therefore, it takes time until the start of using the production service. In this embodiment, as the GA, it is performed with a small number of individuals and generations, but by setting the crossover rate Pc to a high value of 0.9 and searching a wide range, a solution with a certain level of performance is found quickly.

[0168] [Expansion of the directive sentence] In this embodiment, in order to increase the applicable applications, the directive sentence is expanded. Specifically, as the directive sentence for specifying GPU processing, in addition to the kernels directive sentence, it is also expanded to the parallel loop directive sentence and the parallel loop vector directive sentence. In the OpenACC standard, kernels is used for single loop and tightly nested loop. Also, parallel loop is used for loops including non-tightly nested loops. Parallel loop vector is used for loops that cannot be parallelized but can be vectorized. Here, a tightly nested loop is a simple loop in which, in a nested loop, for example, when two loops incrementing i and j are nested, the processing using i and j is performed in the lower loop and not in the upper loop. Also, in the implementation of the PGI compiler etc., there is a difference that the compiler makes the determination of parallelization for kernels, and the programmer makes the determination of parallelization for parallel.

[0169] Therefore, in this embodiment, kernels are used for single and tightly nested loops, and parallel loops are used for non-tightly nested loops. Also, parallel loop vectors are used for loops that cannot be parallelized but can be vectorized. Here, there is a concern that the reliability of the result may be lower than that of the case of kernels by using parallel directive statements. However, it is assumed that a sample test is performed on the final offloaded program, the result difference from the CPU is checked, and the user is shown the result and asked to confirm it. After all, since the hardware is different between the CPU and the GPU, there are differences in the number of significant digits and rounding errors, etc., and it is necessary to check the result difference from the CPU even with only kernels.

[0170] [Settings of resource ratio and resource amount, and arrangement flowchart of new application] FIG. 10 is a flowchart for explaining the settings of the resource ratio and resource amount and the arrangement of a new application added after the GPU offload trial. The flowchart shown in FIG. 10 is executed after the GPU offload trial shown in FIGS. 9A - B.

[0171] In step S51, the resource ratio determination unit 115 acquires the user operation conditions, the CPU processing time of the test case, and the offload device processing time. The user operation conditions are specified by the user when the user specifies the code to be offloaded. The user operation conditions are used when the resource amount setting unit 116 determines the resource amount with reference to the information in the equipment resource DB 132.

[0172] [Processing of resource ratio determination unit 115] In step S52, the resource ratio determination unit 115 determines, based on the performance measurement result, the ratio of the processing times of the CPU and the offload device (the CPU processing time of the test case and the offload device processing time) as the resource ratio.

[0173] With this automatic offloading, during code conversion, the performance measurement results in the verification environment have already been obtained. Using these performance measurement results, the resource ratio determination unit 115 determines the resource ratio between the CPU and the offloading device. Specifically, it determines an appropriate resource ratio based on the ratio of the processing times of the CPU and the offloading device in the verification environment. For example, if the test case processing times in the verification environment are CPU processing: 10 seconds and GPU processing: 5 seconds, the resource ratio is CPU:GPU = 2:1.

[0174] The resource ratio determination unit 115 determines the resource ratio so that the processing times of the CPU and the offloading device are of the same order. By determining the resource ratio so that the processing times of the CPU and the offloading device are of the same order, the processing times of the CPU and the offloading device can be made equal, and the resource amount can be appropriately set even when the CPU and the accelerators are in a mixed environment such as a GPU, an FPGA, or a multi-core CPU.

[0175] When the difference in the processing times of the CPU and the offloading device is equal to or greater than a predetermined threshold, the resource ratio determination unit 115 sets the resource ratio to a predetermined upper limit value. That is, if the difference in the processing times of the CPU and the offloading device in the verification environment is, for example, 10 times or more, setting the resource ratio to 10 times or more will lead to deterioration of cost performance. In this case, for example, the resource ratio such as 5:1 is set as the upper limit (the upper limit value is the resource ratio of 5:1 of the processing time). By setting an upper limit for the resource ratio, a significant increase in the number of VMs can be prevented.

[0176] <Processing of the resource amount setting unit 116> In step S53, the resource amount setting unit 116 sets the resource amount based on the user operation conditions and the appropriate resource ratio. That is, the resource amount setting unit 116 determines the resource amount while keeping the resource ratio as much as possible so as to satisfy the cost conditions specified by the user.

[0177] The resource amount setting unit 116 sets the maximum resource amount that satisfies the user operation conditions while maintaining an appropriate resource ratio. For a specific example, assume that CPU 1VM costs 1000 yen per month, GPU costs 4000 yen per month, and an appropriate resource ratio is 2:1, and the user has a budget within 10000 yen per month. In this case, 2 CPUs and 1 GPU are secured and arranged in a commercial environment.

[0178] If the resource amount setting unit 116 cannot satisfy the user operation conditions even with the minimum resource amount while maintaining the resource ratio, it breaks the resource ratio and sets the minimum resource amounts of the CPU and the offload device so as to satisfy the cost conditions. For a specific example, assume that CPU 1VM costs 1000 yen per month, GPU costs 4000 yen per month, and an appropriate resource ratio is 2:1, and the user has a budget within 5000 yen per month. In this case, since the user budget is insufficient, the resource ratio cannot be maintained, but the resource amounts of the CPU and the offload device are set smaller, that is, 1 CPU and 1 GPU are secured and arranged.

[0179] After finishing the process of step S53 and securing and arranging resources in a commercial environment, before the user uses them, in order to check performance and cost, the automatic verification described in FIG. 2 is executed. As a result, resources can be secured in a commercial environment, and after automatic verification, performance and cost can be presented to the user.

[0180] <Summary of setting resource ratio and resource amount> In order to optimize the resource ratio, the performance measurement results when determining the solution of the offload pattern are used. In implementation, the resource ratio is determined from the processing time of the test case so that the processing times of the CPU and the GPU are of the same order. For example, when the processing time of the test case is CPU processing: 10 seconds and GPU processing: 5 seconds, since the resources on the CPU side are considered to be about twice as much for the same processing time, the resource ratio is 2:1. Note that since the number of virtual machines, etc. is an integer, the resource ratio is rounded to an integer ratio when calculated from the processing time.

[0181] Once the resource ratio is determined, next, the amount of resources to be set when deploying the application to the commercial environment is determined. In implementation, for determining the amount of resources, in order to meet the cost requirements specified by the user at the time of the offloading request, the resource ratio is kept as much as possible, and the number of VMs etc. is determined. Specifically, within the cost range, while keeping the resource ratio, the maximum value is selected for the number of VMs etc.

[0182] For example, regarding the CPU, 1 VM is 1000 yen / month, for the GPU it is 4000 yen / month, and a resource ratio of 2:1 is appropriate. If the user has a budget within 10000 yen per month, then 2 CPUs and 1 GPU are secured. Also, if the resource ratio cannot be kept within the cost range, starting from 1 unit of CPU and 1 unit of GPU, the amount of resources is set so as to be as close as possible to an appropriate resource ratio. For example, if the budget is within 5000 yen per month, the resource ratio cannot be kept, but 1 CPU and 1 GPU are secured. When setting the amount of resources, in implementation, for example, using the virtualization function of Xen Server, resources such as CPUs and GPUs are allocated.

[0183] In step S54, based on the server in the facility resource DB132, the link specification information, and the deployment information of the existing application, the placement setting unit 170 calculates and sets the placement destination (the placement location of the APL) of the new application using the linear programming method.

[0184] [Optimal Placement of the Converted Application] When the offloading server 1 of this embodiment offloads the CPU-oriented program to a device such as a GPU, it optimizes the placement destination so that the application meets the user's cost etc. requirements and operates with a short response time etc.

[0185] [Optimization of the Application Placement Location]< In this embodiment, it is assumed that the application can be deployed not only in the cloud but also at the network edge or user edge. However, compared with the cloud, the server aggregation degree at the network edge or user edge is lower and more dispersed. For this reason, the cost of computing resources is relatively higher than that in the cloud. That is, generally, although the prices of hardware such as CPUs and GPUs are constant regardless of the location of deployment, in a data center operating the cloud, the aggregated servers can be monitored and air-conditioned collectively, etc., so the operating cost is relatively low. For example, as a simple topology of the computing node link, Fig. 11 can be cited.

[0186] Fig. 11 is a diagram showing an example of the topology of computing nodes. Fig. 11 shows a topology in which data is sent from IoT devices, etc. that collect data in the user environment like an IoT system to the user edge, then sent to the cloud via the network edge, and the analysis results are used, for example, by the company's executives to view.

[0187] As shown in Fig. 11, the topology for deploying the application is composed of three layers. The number of sites in the cloud layer (for example, the data center) is "2" (n13, n14), the carrier edge layer (for example, the premises) is "3", the user edge layer (for example, the user environment) is "4" (n6 - n9), and the input nodes are "5" (n1 - n5). Assuming an application such as IoT, IoT data (such as a pollen sensor or a body temperature sensor, which are one of the IoT devices) is collected from the input nodes to the user edge. Depending on the characteristics of the application (such as the required response time conditions), the data is analyzed and processed at the user edge or carrier edge, or the data is sent up to the cloud and then analyzed and processed. The output node is "1" (n15), and the company's executives view the analysis results. For example, when the input node is IoT data (pollen sensor), the statistical and analysis results of the output node are confirmed by the person in charge of the Meteorological Agency. The three-layer placement topology shown in FIG. 11 is an example, and it may be, for example, five layers. Also, the number of user edges and carrier edges may actually be in the dozens to hundreds in some cases.

[0188] The computing nodes can be divided into three types: CPU, GPU, and FPGA. Nodes equipped with GPUs or FPGAs also have CPUs installed, but through virtualization technologies (e.g., NVIDIA vGPU), they are provided in a divided form including CPU resources as GPU instances and FPGA instances.

[0189] Applications are deployed in the cloud, carrier edge, and user edge. The closer to the user environment, the more the response time can be reduced, but the cost of computing resources becomes higher instead. In this embodiment, applications converted for GPUs or FPGAs will be deployed. When deploying, the user can issue two types of requests. The first is the cost requirement, which is to specify the cost of computing resources that can be tolerated to operate the application, for example, to operate within 5000 yen per month. The second is the response time requirement, which is to specify the acceptable response time when operating the application, for example, to return a response within 10 seconds. In conventional facility designs, for example, the location of servers accommodating virtual networks is planned and designed by looking at long-term trends such as traffic increase.

[0190] This embodiment has the following features (1) and (2). (1) The applications to be deployed are not statically determined, but are automatically converted for GPUs or FPGAs, and patterns suitable for the usage form are extracted through actual measurement via GA and the like. Therefore, the code and performance of the applications can change dynamically. (2) It is not sufficient to only reduce the carrier's equipment cost and overall response time. It is necessary to meet individual user requirements for the cost of computing resources and response time. Also, the application placement policy can change dynamically.

[0191] Based on the above features (1) and (2), when there is a placement request from the user, the application placement in this embodiment performs conversion and sequentially places the converted applications on appropriate servers at that time. If the cost performance does not improve even after converting the application, the application placement before conversion is used. For example, when a GPU instance costs twice as much as a CPU instance and the performance is not improved by more than twice even after conversion, it is better to place the application before conversion. Also, when the computing resources and bandwidth have already been used up to the upper limit, it may not be possible to place the application on that server.

[0192] <Linear programming equation for appropriate application placement> In this embodiment, a formulation of the linear programming method is performed to calculate an appropriate placement location for the application. Specifically, the linear programming method uses the parameters of the linear programming equations shown in [Equation 1] (the following equations (1) to (4)) and [Equation 2] (the following equations (3) to (6)).

[0193] Here, the costs of devices and links, the upper limit of computing resources, the upper limit of bandwidth, etc. depend on the servers and networks prepared by the operator. Therefore, the operator sets those parameter values in advance. The amount of computing resources, bandwidth, data volume, and processing time used by the application when offloaded are determined by the measured values in the finally selected offloading pattern in the test in the verification environment before automatic conversion and are automatically set by the environment adaptation function. Whether the user requirement is a cost requirement for computing resources or a response time requirement, the objective function and constraint conditions in the parameters of the linear programming equation change.

[0194] · Parameters of the linear programming equation according to the cost requirement When there is a requirement for placement within a certain amount per month due to the cost requirement, the parameters of the linear programming equation shown in the following [Equation 1] are used.

[0195]

Number

[0196] The minimization of the response time in Equation (1) is the objective function. One of the constraints is the cost of the computing resources in Equation (2) within what limit. Furthermore, the constraint conditions on whether the resource upper limits of the servers in Equations (3) and (4) are exceeded are also added.

[0197] · Parameters of the linear programming equation according to the response time requirement When, according to the response time requirement, it is required that the placement of the application be within a certain number of seconds for the response time, the parameters of the linear programming equation shown in the following [Equation 2] are used.

[0198]

Number

[0199] The minimization of the cost of the computing resources in Equation (5) corresponding to Equation (2) is the objective function. One of the constraints is the response time in Equation (6) corresponding to Equation (1) within what number of seconds. Furthermore, the constraint conditions in Equations (3) and (4) are also added.

[0200] · Explanation of the parameters of the linear programming equation Equations (1) and (6) are equations for calculating the response time of application k. In the case of Equation (1), Rk is the objective function, and in the case of Equation (6), Rk is a constraint condition for setting the upper limit specified by the user.

[0201] Equations (2) and (5) are equations for calculating the cost (price) Pk for operating application k. In the case of Equation (2), Pk is a constraint condition for setting the upper limit specified by the user, and in the case of Equation (5), Pk is the objective function.

[0202] Equations (3) and (4) are constraint conditions for setting the upper limits of computing resources and communication bandwidth, which are calculated including applications placed by others to prevent exceeding the resource upper limits due to the placement of new user applications.

[0203] By solving the linear programming equations of formulas (1) to (4) and formulas (3) to (6) using a linear programming solver such as GLPK (Gnu Linear Programming Kit) or CPLEX (IBM Decision Optimization) for different conditions of network topology, conversion application type (such as cost increase and performance increase for CPU), user requirements, and existing deployed applications, an appropriate application placement can be calculated. After calculating the appropriate placement, the actual placement is sequentially performed for multiple users, so that multiple applications are placed based on the requirements of each user.

[0204] As described above, when there is a new application placement request, it can be calculated based on the linear programming equation and placed in order, so that a placement that meets user desires is possible. Here, although it can be said that the placement of application programs is a first-come, first-served basis as it is performed sequentially, the proper placement of the already deployed application program group is recalculated periodically, such as every 100 applications. Then, according to the cost and response time specified by the user, a placement that minimizes the objective function can be calculated, and the application can be relocated to the position determined by the calculation.

[0205] [Evaluation] Based on the linear programming equation, which is an aspect of the linear programming method, using the free solver GLPK (registered trademark), it was confirmed by changing several conditions that multiple applications are properly placed.

[0206] <Evaluation Conditions> · Target Application The application to be placed is for image processing by Fourier transform, which is assumed to be used by many users. The Fourier transform process (FFT) is used in various scenarios of IoT monitoring, such as vibration frequency analysis. NAS.FT (https: / / www.nas.nasa.gov / publications / npb.html) (registered trademark) is one of the open-source applications for FFT processing. It performs calculations of the size 2048×2048 for the provided sample tests. When assuming an application for transferring data from a device to a network in IoT, in order to reduce network costs, it is assumed that primary analysis such as FFT processing is performed on the device side and then sent.

[0207] MRI-Q (http: / / impact.crhc.illinois.edu / parboil / ) (registered trademark) calculates the matrix Q representing the scanner configuration for calibration used in the 3D MRI reconstruction algorithm in non-Cartesian space. In the IoT environment, image processing is often required for automatic monitoring from camera videos, and there is a need for automatic offloading of image processing. MRI-Q is a C language application that executes 3D MRI image processing during performance measurement and measures the processing time using sample data of the size 64×64×64 for Large. CPU processing is in C language, and FPGA processing is based on OpenCL (registered trademark). With the GPU and FPGA automatic offloading technology of this embodiment, NAS.FT can be accelerated by the GPU, and MRI-Q can be accelerated by the FPGA, achieving a 5-fold and 7-fold acceleration respectively compared to the CPU.

[0208] · Evaluation method The topology for placing the application is composed of three layers as shown in FIG. 11. The number of base points in the cloud layer is "5", the carrier edge layer is "20", the user edge layer is "60", and the input nodes are "300". Assuming applications such as IoT, IoT data, etc. are collected from the input nodes to the user edge, and according to the characteristics of the application (requirements such as response time requirements), analysis and processing are performed at the user edge and carrier edge, or the data is sent up to the cloud and then analyzed and processed.

[0209] Based on the parameters of the linear programming equations shown in [Equation 1] and [Equation 2], and according to the user's requirements, for example, 1000 applications are arranged. The applications are assumed to be IoT applications that analyze data generated from input nodes. A placement request is randomly generated from an input node (assumed to be "300"). For example, as the number of placement requests, the applications are requested to be placed 1000 times at a ratio of NAS.FT:MRI-Q = 3:1. Also, as user requirements, when making a placement request, either a price condition or a response time condition is selected for each application. In the case of NAS.FT, for the price, an upper limit of 7000 yen per month, 8500 yen per month, or 10000 yen per month is selected, and for the response time, an upper limit of 6 seconds, 7 seconds, or 10 seconds is selected. In the case of MRI-Q, for the price, an upper limit of 12500 yen per month or 20000 yen per month is selected, and for the response time, an upper limit of 4 seconds or 8 seconds is selected.

[0210] There are three patterns as variations of the user requirements. Pattern 1: For NAS.FT, select 6 types of requests one-sixth each, and for MRI-Q, select 4 types of requests one-fourth each. Pattern 2: The requests select the condition with the lowest price as the upper limit (initially 7000 yen, 12500 yen), and if there is no availability, use the next lower price condition. Pattern 3: The requests select the condition with the lowest response time as the upper limit (initially 6 seconds, 4 seconds), and if there is no availability, use the next faster response time condition.

[0211] · Simulation of placement The configuration is carried out through simulation experiments using the solver GLPK 5.0 (registered trademark) as an evaluation tool. For the simulation of large-scale network configurations, it will be a simulation using the evaluation tool. In actual use, when an offloading request for an application arrives, an offloading pattern is created through repeated performance tests using the verification environment, and an appropriate amount of resources is determined based on the performance test results in the verification environment (see Figure 10). Then, an appropriate configuration is determined using GLPK or the like according to user requirements, and normal confirmation tests and performance tests when actually deployed are automatically performed. The results and prices are presented to the user, and the use is started after the user makes a decision.

[0212] Figure 12 is a graph showing the change in the number of application configurations with respect to the average response time. Figure 12 shows the average response time and the number of application configurations for the above three patterns. It was confirmed that in Pattern 2, it fills up in order from the cloud, and in Pattern 3, it fills up in order from the edge. In Pattern 1, when various requests arrive, it is arranged to meet the user's requirement conditions. As shown in Figure 12, in Pattern 2, up to around 400 configurations, all are arranged in the cloud and the average response time remains the slowest, but it can be seen that it gradually decreases when the cloud is filled up. In Pattern 3, NAS.FT is arranged from the user edge, and MRI-Q is arranged from the carrier edge. Therefore, the average response time is the shortest. However, as the number increases, it is also arranged in the cloud, so the average response time becomes slower. In Pattern 2, the average response time is in the middle between Pattern 1 and Pattern 3 and is arranged according to user requirements. Therefore, in Pattern 2, the average response time is appropriately reduced compared to Pattern 1 where all initially go into the cloud.

[0213] In this way, when the software is automatically adapted to the deployment destination environment and automatically offloaded to a GPU or the like, it meets the user's cost requirements and response time requirements. That is, the program is converted so that it can be processed by a device such as a GPU, and after the amount of resources to be assigned is determined, the optimal configuration of the converted application is performed.

[0214] To summarize, first, the application's usage data volume, computing resource volume, bandwidth, and processing time are set from the performance test data conducted in the verification environment when converting the program. Based on the values set for each conversion application and the values such as the costs of pre-set servers and links, an appropriate placement of the application is calculated based on a linear programming formula. When placing the application, based on the price and response time requests specified by the user, one becomes the constraint condition and the other becomes the objective function. An appropriate placement is calculated by a linear programming solver, and the proposed method presents the price, etc. to the user when resources are placed at the calculated location, and utilization starts after user approval.

[0215] For applications automatically offloaded to GPUs and FPGAs, the appropriate placement is calculated by changing the price conditions, response time conditions, number of application placements, etc. requested by the user. This enables placement according to user requirements.

[0216] (Second Embodiment) Next, the offload server 1A, etc. in the second embodiment of the present invention will be described. The second embodiment is an example applied to the automatic offloading of loop statements to FPGAs. This embodiment will describe an example applied to an FPGA (Field Programmable Gate Array) as a PLD (Programmable Logic Device). The present invention is applicable to programmable logic devices in general.

[0217] (Principle Explanation) Since it is difficult to predict which loop should be offloaded by the FPGA to achieve high speed, it is proposed to automatically measure in the same verification environment as the GPU. However, since it takes several hours or more to compile OpenCL and run it on the actual machine for the FPGA, it is impossible to perform repeated measurements using GA with automatic GPU offloading because the processing time would be enormous. Therefore, the measurement is performed after narrowing down the candidate loop statements to be offloaded to the FPGA. Specifically, for the discovered loop statements, arithmetic intensity analysis tools such as ROSE (registered trademark) are used to extract the loop statements with high arithmetic intensity. Furthermore, profiling tools such as gcov (registered trademark) are used to extract the loop statements with a large number of loop iterations.

[0218] The loop statements with high arithmetic intensity and a large number of loop iterations are used as candidates for OpenCL conversion. When converting to OpenCL, the CPU processing program is divided into a kernel (FPGA) and a host (CPU) according to the grammar of OpenCL. For the candidate loop statements, the created OpenCL is pre-compiled to find the loop statements with high resource efficiency. Since the resources to be created are known during compilation, the loop statements with a sufficiently small amount of resources to be used are further narrowed down. Since several candidate loop statements remain, the performance and power consumption are actually measured using them. The selected single loop statement is compiled and measured, and for the single loop statement that can be further accelerated, its combination pattern is also created and measured for the second time. Among the measured multiple patterns, the pattern with short time and low power consumption is selected as the solution.

[0219] Regarding the FPGA offloading of loop statements, after narrowing down using arithmetic intensity etc., measurement is performed, and by increasing the evaluation value of the low-power pattern, automatic high-speed and low-power operation are achieved.

[0220] FIG. 13 is a functional block diagram showing a configuration example of the offloading server 1A according to the second embodiment of the present invention. In the description of this embodiment, the same components as those in FIG. 1 are denoted by the same reference numerals, and the description of overlapping parts is omitted. The offloading server 1A is a device that automatically offloads specific processing of an application to an accelerator. Also, the offloading server 1A can be connected to an emulator. As shown in FIG. 13, the offloading server 1A includes a control unit 21, an input / output unit 12, a storage unit 13, and a verification machine 14 (Verification machine) (device for accelerator verification).

[0221] The control unit 21 is an automatic offloading function unit that controls the entire offloading server 1A. The control unit 21 is realized, for example, by a CPU (not shown) expanding and executing a program (offloading program) stored in the storage unit 13 in the RAM.

[0222] The control unit 21 includes an application code specifying unit (Specify application code) 111, an application code analyzing unit (Analyze application code) 112, a PLD processing specifying unit 213, an arithmetic intensity calculating unit 214, an arrangement setting unit 170, a PLD processing pattern creating unit 215, a performance measuring unit 118, an execution file creating unit 119, a production environment deployment unit (Deploy final binary files to production environment) 120, a performance measurement test extraction execution unit (Extract performance test cases and run automatically) 121, and a user providing unit (Provide price and performance to a user to judge) 122.

[0223] <PLD processing specifying unit 213> The PLD processing specifying unit 213 identifies the loop statements (repetitive statements) of the application, and creates and compiles a plurality of offloading processing patterns that specify pipeline processing and parallel processing in the PLD for each identified loop statement in OpenCL. The PLD processing specification unit 213 includes an offload range extraction unit (Extract offload able area) 213a and an intermediate language file output unit (Output intermediate file) 213b.

[0224] The offload range extraction unit 213a identifies processing that can be offloaded to the FPGA, such as loop statements and FFT, and extracts intermediate language corresponding to the offload processing.

[0225] The intermediate language file output unit 213b outputs the extracted intermediate language file 133. The intermediate language extraction is not a one-time process but is repeated for appropriate offload area exploration and for trial execution and optimization.

[0226] <Arithmetic intensity calculation unit 214> The arithmetic intensity calculation unit 214 calculates the arithmetic intensity of the loop statements of the application using an arithmetic intensity analysis tool such as the ROSE framework (registered trademark). The arithmetic intensity is the value obtained by dividing the number of floating point operations (floating point number, FN) executed during the operation of the program by the number of bytes accessed to the main memory (FN operations / memory access). The arithmetic intensity is an index that increases when the number of calculations is large and decreases when the number of accesses is large. Processing with a high arithmetic intensity is heavy processing for the processor. Therefore, an arithmetic intensity analysis tool is used to analyze the arithmetic intensity of the loop statements. The PLD processing pattern creation unit 215 narrows down the loop statements with a high arithmetic intensity as offload candidates.

[0227] An example of the calculation of arithmetic intensity will be described. Assume that 10 floating point calculations (10 FLOP) are performed in one loop, and the data used in the loop is 2 bytes. When data of the same size is used for each loop, 10 / 2 = 5 [FLOP / byte] becomes the arithmetic intensity. Note that since the loop count is not considered in the arithmetic intensity, in this embodiment, in addition to the arithmetic intensity, the loop count is also considered for narrowing down.

[0228] <PLD processing pattern creation unit 215> Based on the arithmetic intensity calculated by the arithmetic intensity calculation unit 214, the PLD processing pattern creation unit 215 narrows down loop statements with an arithmetic intensity higher than a predetermined threshold (hereinafter, appropriately referred to as high arithmetic intensity) as offload candidates, and creates a PLD processing pattern. Also, as a basic operation, the PLD processing pattern creation unit 215 creates a PLD processing pattern that excludes loop statements (iteration statements) that result in compilation errors from being offloaded and specifies whether to perform PLD processing on iteration statements that do not result in compilation errors.

[0229] · Loop count measurement function As a loop count measurement function, the PLD processing pattern creation unit 215 measures the loop count of the loop statements of the application using a profiling tool, and narrows down loop statements with high arithmetic intensity and a loop count greater than a predetermined number (hereinafter, appropriately referred to as high loop count). To grasp the loop count, gcov of GNU coverage, etc. are used. As profiling tools, "GNU profiler (gprof)" and "GNU coverage (gcov)" are known. Since both can investigate the number of executions of each loop, either can be used.

[0230] Also, in arithmetic intensity analysis, since the loop count is not particularly visible, in order to detect loops with a large loop count and high load, the loop count is measured using a profiling tool. Here, the height of the arithmetic intensity indicates whether the processing is suitable for offloading to the FPGA, and the loop count × arithmetic intensity indicates whether the load related to offloading to the FPGA is high.

[0231] · OpenCL (intermediate language) creation function The PLD processing pattern creation unit 215 creates (OpenCLizes) OpenCL for offloading each narrowed-down loop statement to the FPGA as an OpenCL creation function. That is, the PLD processing pattern creation unit 215 compiles the OpenCL for offloading the narrowed-down loop statement. Also, the PLD processing pattern creation unit 215 lists the loop statements that have been performance-measured and have been made to have higher performance than the CPU, and creates OpenCL for offloading by combining the loop statements in the list.

[0232] Describe OpenCLization. The PLD processing pattern creation unit 215 converts the loop statement into a high-level language such as OpenCL. First, the program for CPU processing is divided into a kernel (FPGA) and a host (CPU) according to the grammar of a high-level language such as OpenCL. For example, when one of the 10 for statements is to be processed by the FPGA, that one is cut out as a kernel program and described according to the grammar of OpenCL. Examples of the grammar of OpenCL will be described later.

[0233] Furthermore, techniques for further speeding up can be incorporated when dividing. Generally, in order to speed up using an FPGA, there are local memory cache, stream processing, multiple instantiation, loop statement unfolding processing, nested loop statement integration, memory interleaving, etc. Although these do not necessarily have an absolute effect depending on the loop statement, they are often used as techniques for speeding up.

[0234] The kernel created according to the grammar of the C language of OpenCL is executed on the device (for example, FPGA) by the program on the host (for example, CPU) side that creates it using the runtime API of the C language of OpenCL. The part that calls the kernel function hello() from the host side is to call clEnqueueTask(), which is one of the OpenCL runtime APIs. The basic flow of initializing, executing, and terminating OpenCL described in the host code is as follows in steps 1 to 13 below. Among these steps 1 to 13, steps 1 to 10 are the procedures (preparations) until the kernel function hello() is called from the host side, and the kernel is executed in step 11.

[0235] 1. Platform identification Use the function clGetPlatformIDs() that provides platform-specific functions defined in the OpenCL runtime API to identify the platform on which OpenCL operates.

[0236] 2. Device identification Use the function clGetDeviceIDs() that provides device-specific functions defined in the OpenCL runtime API to identify devices such as GPUs used on the platform.

[0237] 3. Context creation Use the function clCreateContext() that provides context creation functions defined in the OpenCL runtime API to create an OpenCL context that is the execution environment for running OpenCL.

[0238] 4. Command queue creation Use the function clCreateCommandQueue() that provides command queue creation functions defined in the OpenCL runtime API to create a command queue which is preparation for controlling the device. In OpenCL, through the command queue, operations on the device from the host (issuing kernel execution commands and host-device memory copy commands) are executed.

[0239] 5. Memory object creation Use the function clCreateBuffer() that provides functions for allocating memory on the device defined in the OpenCL runtime API to create a memory object that allows the host side to reference the memory object.

[0240] 6. Kernel File Loading The kernel to be executed on the device is controlled by a host-side program during its execution. Therefore, the host program first needs to load the kernel program. The kernel program includes binary data created by the OpenCL compiler and source code written in OpenCL C language. This kernel file is loaded (description omitted). Note that the OpenCL runtime API is not used for kernel file loading.

[0241] 7. Program Object Creation In OpenCL, the kernel program is recognized as a program project. This procedure is program object creation. Using the function clCreateProgramWithSource() that provides the program object creation function defined in the OpenCL runtime API, create a program object that allows the host side to reference the memory object. When creating from the compiled binary sequence of the kernel program, use clCreateProgramWithBinary().

[0242] 8. Build Build the program object using the OpenCL C compiler and linker registered as source code. Using the function clBuildProgram() that executes the build by the OpenCL C compiler and linker defined in the OpenCL runtime API, build the program object. Note that this compilation procedure is not required when generating a program object from the compiled binary sequence with clCreateProgramWithBinary().

[0243] 9. Kernel Object Creation Create a kernel object using the function `clCreateKernel()`, which provides the kernel object creation function defined in the OpenCL runtime API. Since one kernel object corresponds to one kernel function, specify the name of the kernel function (`hello`) when creating the kernel object. Also, when multiple kernel functions are described as one program object, since one kernel object corresponds one-to-one to one kernel function, call `clCreateKernel()` multiple times.

[0244] 10. Setting Kernel Arguments Set kernel arguments using the function `clSetKernel()`, which provides the function to give arguments to a kernel (pass values to the arguments of the kernel function) defined in the OpenCL runtime API. With the above preparations completed in steps 1 to 10, we enter step 11 of executing the kernel on the device from the host side.

[0245] 11. Kernel Execution Kernel execution (submitted to the command queue) acts on the device, so it is the queuing function for the command queue. Queue a command to execute the kernel `hello` on the device using the function `clEnqueueTask()`, which provides the kernel execution function defined in the OpenCL runtime API. After the command to execute the kernel `hello` is queued, it will be executed on the executable arithmetic units on the device.

[0246] 12. Reading from Memory Objects Using the function `clEnqueueReadBuffer()` that provides the function of copying data from the device-side memory defined in the OpenCL runtime API to the host-side memory, data is copied from the device-side memory area to the host-side memory area. Also, using the function `clEnqueueWriteBuffer()` that provides the function of copying data from the host-side to the client-side memory, data is copied from the host-side memory area to the device-side memory area. Note that since these functions act on the device, the data copy will start after the copy command is queued to the command queue once.

[0247] 13. Object Release Finally, release various objects created so far. The above described the device execution of the kernel created along the OpenCL C language.

[0248] · Resource Amount Calculation Function The PLD processing pattern creation unit 215 calculates the resource amount to be used by pre-compiling the created OpenCL as a resource amount calculation function ("first resource amount calculation"). The PLD processing pattern creation unit 215 calculates the resource efficiency based on the calculated arithmetic intensity and resource amount, and selects c loop statements with a resource efficiency higher than a predetermined value in each loop statement based on the calculated resource efficiency. The PLD processing pattern creation unit 215 calculates the resource amount to be used by pre-compiling the combined offload OpenCL ("second resource amount calculation"). Here, it may be the sum of the resource amounts in the pre-compilation before the first measurement without pre-compiling.

[0249] <Performance Measurement Unit 118> The performance measurement unit 118 compiles the application of the created PLD processing pattern, arranges it on the verification machine 14, and executes the performance measurement processing when offloaded to the PLD.

[0250] The performance measurement unit 118 executes the arranged binary file, measures the performance when offloading, and returns the performance measurement result to the offloading range extraction unit 213a. In this case, the offloading range extraction unit 213a performs extraction of another PLD processing pattern, and the intermediate language file output unit 213b attempts performance measurement based on the extracted intermediate language (see reference sign a in FIG. 2).

[0251] The performance measurement unit 118 includes a binary file deployment unit 118a. The binary file deployment unit 118a deploys (arranges) an executable file derived from the intermediate language to the verification machine 14 equipped with a GPU.

[0252] A specific example of performance measurement will be described. The PLD processing pattern creation unit 215 narrows down high resource - efficient loop statements and compiles OpenCL for offloading the loop statements narrowed down by the executable file creation unit 119. The performance measurement unit 118 measures the performance of the compiled program ("first - time performance measurement").

[0253] Then, the PLD processing pattern creation unit 215 lists loop statements that have been measured to have higher performance compared to the CPU. The PLD processing pattern creation unit 215 creates OpenCL for offloading by combining the loop statements in the list. The PLD processing pattern creation unit 215 calculates the amount of resources to be used by pre - compiling with the combined offloading OpenCL. Note that it may also be the sum of the resource amounts in the pre - compilation before the first - time measurement without pre - compilation. The executable file creation unit 119 compiles the combined offloading OpenCL, and the performance measurement unit 118 measures the performance of the compiled program ("second - time performance measurement").

[0254] <Executable file creation unit 119> The executable file creation unit 119 selects the PLD processing pattern with the highest evaluation value from a plurality of PLD processing patterns based on the measurement results of the processing time repeated a predetermined number of times, and compiles the PLD processing pattern with the highest evaluation value to create an executable file.

[0255] Hereinafter, the automatic offloading operation of the offloading server 1A configured as described above will be described. [Automatic offloading operation] The offloading server 1A of the present embodiment is an example applied to the FPGA automatic offloading of user application logic as an elemental technology of environment-adaptive software. The automatic offloading process of the offloading server 1A shown in FIG. 2 will be described with reference thereto. As shown in FIG. 2, the offloading server 1A is applied to the elemental technology of environment-adaptive software. The offloading server 1A includes a control unit (automatic offloading function unit) 11, a test case DB 131, an intermediate language file 133, and a verification machine 14. The offloading server 1A acquires the application code 125 used by the user.

[0256] The user uses, for example, various devices 151, a device 152 having a CPU-GPU, a device 153 having a CPU-FPGA, and a device 154 having a CPU. The offloading server 1A automatically offloads the functional processing to the accelerators of the device 152 having a CPU-GPU and the device 153 having a CPU-FPGA.

[0257] Hereinafter, the operations of each unit will be described with reference to the step numbers in FIG. 2. <Step S21: Specify application code> In step S21, the application code specifying unit 111 (see FIG. 13) specifies the processing function (such as image analysis) of the service provided to the user. Specifically, the application code specifying unit 111 specifies the input application code.

[0258] <Step S12: Analyze application code> In step S12, the application code analysis unit 112 (see FIG. 13) analyzes the source code of the processing function and grasps the structure of using specific libraries such as loop statements and FFT library calls.

[0259] <Step S13: Extract offloadable area> In step S13, the PLD process specification unit 213 (see FIG. 13) identifies the loop statements (repetitive statements) of the application, specifies parallel processing or pipeline processing in the FPGA for each repetitive statement, and compiles them with a high-level synthesis tool. Specifically, the offload range extraction unit 213a (see FIG. 13) identifies the processing that can be offloaded to the FPGA, such as loop statements, and extracts OpenCL as the intermediate language corresponding to the offload process.

[0260] <Step S14: Output intermediate file> In step S14, the intermediate language file output unit 213b (see FIG. 13) outputs the intermediate language file 133. The extraction of the intermediate language is not completed at once, and is repeated for trial execution and optimization for appropriate offload area exploration.

[0261] <Step S15: Compile error> In step S15, the PLD process pattern creation unit 215 (see FIG. 13) creates a PLD process pattern that excludes the loop statements with compilation errors from the offload targets and specifies whether to perform FPGA processing or not for the repetitive statements without compilation errors.

[0262] <Step S21: Deploy binary files> In step S21, the binary file placement unit 118a (see FIG. 13) deploys an executable file derived from the intermediate language to the verification machine 14 equipped with an FPGA. The binary file placement unit 118a starts the placed file, executes the assumed test case, and measures the performance when offloaded.

[0263] <Step S22: Measure performances> In step S22, the performance measurement unit 118 (see FIG. 13) executes the placed file and measures the performance and power consumption when offloaded. To make the offloaded area more appropriate, this performance measurement result is returned to the offload range extraction unit 213a, and the offload range extraction unit 213a performs extraction of another pattern. Then, the intermediate language file output unit 213b attempts performance measurement based on the extracted intermediate language (see reference sign a in FIG. 2). The performance measurement unit 118 repeatedly measures the performance and power consumption in the verification environment, and finally determines the code pattern to be deployed.

[0264] As shown by reference sign a in FIG. 2, the control unit 21 repeatedly executes the above steps S12 to S22. Summarizing the automatic offload function of the control unit 21, it is as follows. That is, the PLD process designation unit 213 identifies the loop statements (repetition statements) of the application, and for each repetition statement, specifies parallel processing or pipeline processing in the FPGA in OpenCL (intermediate language), and compiles it with a high-level synthesis tool. Then, the PLD process pattern creation unit 215 sets the loop statements that cause compilation errors as not being offload targets, and creates a PLD process pattern for specifying whether to perform PLD processing on the loop statements that do not cause compilation errors. Then, the binary file placement unit 118a compiles the application corresponding to the PLD process pattern and places it in the verification machine 14, and the performance measurement unit 118 executes the performance measurement process on the verification machine 14. The executable file creation unit 119, based on the performance measurement results repeated a predetermined number of times, selects the highest evaluation value from a plurality of PLD process patterns (for example, evaluation value = (processing time) -1 / 2Select the pattern with the highest (value), compile the selected pattern, and create an executable file.

[0265] <Step S23: Deploy final binary files to production environment> In step S23, the production environment deployment unit 120 determines the pattern that designates the final offload area and deploys it to the production environment for users.

[0266] <Step S24: Extract performance test cases and run automatically> In step S24, after the executable file is deployed, the performance measurement test extraction and execution unit 121 extracts performance test items from the test case DB131 to show the performance to the user and automatically executes the extracted performance tests.

[0267] <Step S25: Provide price and performance to a user to judge> In step S25, the user provision unit 122 presents information such as price and performance to the user based on the performance test results. The user judges whether to start using and paying for the service based on the presented information such as price and performance.

[0268] The above steps S21 to S25 are performed in the background of the user's service usage, for example, it is assumed that they are performed during the first day of trial use. Also, the processing performed in the background for cost reduction may target only GPU·FPGA offloading.

[0269] As described above, when applied to the elemental technologies of the environment-adaptive software, the control unit (automatic offloading function unit) 21 of the offloading server 1A extracts an offloading area from the source code of the application used by the user and outputs an intermediate language for offloading of functional processing (steps S12 to S15). The control unit 21 arranges and executes the executable file derived from the intermediate language on the verification machine 14 to verify the offloading effect (steps S21 to S22). After repeating the verification and determining an appropriate offloading area, the control unit 21 deploys the executable file to the production environment actually provided to the user and provides it as a service (step S26).

[0270] In the above, the processing flow for collectively performing code conversion, resource amount adjustment, and placement location adjustment, which are necessary for environment adaptation, has been described. However, the present invention is not limited to this, and it is also possible to cut out only the processing to be performed. For example, when it is desired to perform only code conversion for an FPGA, only the necessary parts such as the environment adaptation function and verification environment in steps S21 to S25 described above may be used.

[0271] [FPGA Automatic Offloading] The above-described code analysis analyzes the application code using a syntax analysis tool such as Clang. Since code analysis requires analysis assuming the device to be offloaded, generalization is difficult. However, it is possible to grasp the structure of the code such as loop statements and variable reference relationships, or to grasp that it is a functional block that performs FFT processing as a functional block, or that a library that performs FFT processing is called. It is difficult for the offloading server to automatically determine the functional block. It is also possible to grasp this by similarity determination or the like using a similar code detection tool such as Deckard. Here, Clang is a tool for C / C++, but it is necessary to select a tool according to the language to be analyzed.

[0272] Also, when offloading the processing of an application, it is necessary to consider each of the GPU, FPGA, IoT GW, etc. according to the offloading destination. Generally, it is difficult to automatically discover the maximum performance setting at once regarding performance. Therefore, the offloading pattern is tried by repeating performance measurement several times in the verification environment to find a pattern that can achieve speedup.

[0273] Hereinafter, an FPGA offloading method for the loop statement of application software will be described. [Flowchart] FIG. 14 is a flowchart for explaining the outline of the operation of the offloading server 1A. In step S201, the application code analysis unit 112 analyzes the source code of the application to be offloaded. The application code analysis unit 112 analyzes information on loop statements and variables according to the language of the source code.

[0274] In step S202, the PLD processing specification unit 213 identifies the loop statement and reference relationship of the application.

[0275] Next, the PLD processing pattern creation unit 215 performs a process of narrowing down candidates for whether to try FPGA offloading for the grasped loop statement. Whether there is an offloading effect on the loop statement is indicated by arithmetic intensity. In step S203, the arithmetic intensity calculation unit 214 calculates the arithmetic intensity of the loop statement of the application using an arithmetic intensity analysis tool. Arithmetic intensity is an index that increases when the number of calculations is large and decreases when the number of accesses is large. A process with a high arithmetic intensity is a heavy process for the processor. Therefore, the arithmetic intensity analysis tool analyzes the arithmetic intensity of the loop statement and narrows down the loop statements with high density as offloading candidates. Therefore, the arithmetic intensity analysis tool analyzes the arithmetic intensity of the loop statement and narrows down the loop statements with high density as offloading candidates.

[0276] Even for a loop statement with high arithmetic intensity, when processing it on an FPGA, excessive consumption of FPGA resources is a problem. Therefore, the calculation of the amount of resources when processing a high arithmetic intensity loop statement on an FPGA will be described. As a process when compiling for an FPGA, it is converted from a high-level language such as OpenCL to the level of hardware description such as HDL, and based on that, actual wiring processing and the like are performed. At this time, the wiring processing and the like take a long time, but up to the intermediate state stage of HDL and the like, the time is only in minutes. Even at the intermediate state stage of HDL and the like, the resources such as Flip Flop and Look Up Table used in the FPGA are known. Therefore, looking at the intermediate state stage of HDL and the like, the amount of resources used can be known in a short time even before the compilation is completed.

[0277] Therefore, in this embodiment, the PLD processing pattern creation unit 215 first converts the target loop statement into a high-level language such as OpenCL and calculates the amount of resources. Also, since the arithmetic intensity and the amount of resources are determined when the loop statement is offloaded, the arithmetic intensity / resource amount or arithmetic intensity × loop count / resource amount is defined as the resource efficiency. Then, the loop statements with high resource efficiency are further narrowed down as offload candidates.

[0278] Returning to the flow of FIG. 14, in step S204, the PLD processing pattern creation unit 215 measures the loop count of the loop statements of the application using a profiling tool such as gcov or gprof. In step S205, the PLD processing pattern creation unit 215 narrows down the loop statements with high arithmetic intensity and high loop count among the loop statements.

[0279] In step S206, the PLD processing pattern creation unit 215 creates OpenCL for offloading each of the narrowed-down loop statements to the FPGA.

[0280] Here, a supplementary explanation will be given regarding the OpenCL conversion (creation of OpenCL) of loop statements. That is, when converting a loop statement into a high-level language such as OpenCL, two processes are required. One is to divide the program of CPU processing into a kernel (FPGA) and a host (CPU) according to the grammar of a high-level language such as OpenCL. The other is to incorporate techniques for speeding up during the division. Generally, in order to speed up using an FPGA, there are local memory caches, stream processing, multiple instantiations, loop statement unfolding processing, nested loop statement integration, memory interleaving, etc. These are not always absolutely effective depending on the loop statement, but are often used as techniques for speeding up.

[0281] Next, since several loop statements with high resource efficiency are selected, create as many offload patterns for actually measuring the performance using them as the number of times for actual measurement. The speedup with an FPGA can be in a form where the FPGA resource amount is intensively applied to only one process for speedup, or in a form where the FPGA resources are distributed among multiple processes for speedup. Create a certain number of patterns of the selected single loop statement and pre-compile them as a pre-stage for operating on an actual FPGA device.

[0282] In step S207, the PLD processing pattern creation unit 215 calculates the resource amount to be used by pre-compiling the created OpenCL ("first resource amount calculation").

[0283] In step S208, the PLD processing pattern creation unit 215 narrows down the loop statements with high resource efficiency.

[0284] In step S209, the execution file creation unit 119 compiles the OpenCL for offloading the narrowed-down loop statements.

[0285] In step S210, the performance measurement unit 118 measures the performance of the compiled program ("first performance measurement"). Since there are still some candidate loop statements, the performance measurement unit 118 actually measures the performance using them (for details, refer to the subroutine in FIG. 15).

[0286] In step S211, the PLD processing pattern creation unit 215 lists the loop statements that have been measured for performance and have been enhanced in performance compared to the CPU.

[0287] In step S212, the PLD processing pattern creation unit 215 creates OpenCL for offloading by combining the loop statements in the list. In step S213, the PLD processing pattern creation unit 215 calculates the amount of resources to be used by pre-compiling with the combined offloading OpenCL ("second resource amount calculation"). Note that instead of pre-compiling, the sum of the resource amounts in the pre-compilation before the first measurement may be used. In this way, the number of pre-compilation times can be reduced.

[0288] In step S214, the execution file creation unit 119 compiles the combined offloading OpenCL.

[0289] In step S215, the performance measurement unit 118 measures the performance of the compiled program ("second performance measurement"). The performance measurement unit 118 compiles and measures for the selected single loop statement, and for the single loop statement that can be further accelerated, its combination pattern is also created and the second performance measurement is performed (for details, refer to the subroutine in FIG. 15).

[0290] In step S216, the production environment deployment unit 120 selects the pattern with the highest performance among the first and second measurements and ends the processing of this flow. Among the measured multiple patterns, the pattern with a short time is selected as the solution.

[0291] In this way, the FPGA automatic offloading of loop statements focuses on loop statements with high arithmetic intensity, high number of loop iterations, and high resource efficiency, creates an offloading pattern, and performs high-speed pattern search through actual measurement in the verification environment (see Fig. 14).

[0292] Fig. 15 is a flowchart showing the performance and power consumption measurement process of the performance measurement unit 118. This flow is called and executed by the subroutine call in step S210 or step S215 of Fig. 14.

[0293] In step S301, the performance measurement unit 118 measures the processing time required during FPGA offloading.

[0294] In step S302, the performance measurement unit 118 sets an evaluation value based on the measured processing time.

[0295] In step S303, the performance measurement unit 118 measures the performance of patterns with high evaluation values evaluated so that the higher the evaluation value of an individual, the higher the fitness, and returns to step S210 or step S215 of Fig. 14.

[0296] [Example of Creating Offloading Pattern] Fig. 16 is a diagram showing the search image of the PLD processing pattern creation unit 215. The control unit (automatic offloading function unit) 21 (see Fig. 13) analyzes the application code 125 (see Fig. 2) used by the user, and as shown in Fig. 16, checks the parallelizability of the for statements from the code patterns 241 of the application code 125. As indicated by the symbol r in Fig. 16, when four for statements are found from the code pattern 241, one digit is assigned to each for statement, here four digits of 1 or 0 are assigned to the four for statements. Here, 1 is used when processing by FPGA, and 0 is used when not processing by FPGA (i.e., processing by CPU).

[0297] [Flow from C Code to Search for Final OpenCL Solution] The procedures A - F in FIG. 17 are diagrams explaining the flow from C code to the search for the final OpenCL solution. The application code analysis unit 112 (see FIG. 13) syntax - analyzes the "C code" shown in procedure A of FIG. 17 (<syntax analysis>: see reference symbol s in FIG. 17), and the PLD processing specification unit 213 (see FIG. 13) identifies the "loop statements, variable information" shown in procedure B of FIG. 17 (see reference symbol t in FIG. 17).

[0298] The arithmetic intensity calculation unit 214 (see FIG. 13) performs arithmetic intensity analysis (Arithmetic Intensity analysis) on the identified "loop statements, variable information" using an arithmetic intensity analysis tool (see reference symbol u in FIG. 17). The PLD processing pattern creation unit 215 narrows down the loop statements with high arithmetic intensity as offload candidates. Further, the PLD processing pattern creation unit 215 performs profiling analysis (Profiling analysis) using a profiling tool to further narrow down the loop statements with high arithmetic intensity and high loop counts.

[0299] Then, the PLD processing pattern creation unit 215 creates (OpenCL - converts) OpenCL for offloading each narrowed - down loop statement to the FPGA (see reference symbol v in FIG. 17). Furthermore, high - speed techniques such as unfolding are introduced together with code splitting during OpenCL conversion (described later).

[0300] <Specific example of "high arithmetic intensity, OpenCL conversion" (Part 1): Procedure C> For example, when four for loops (assignment of 4 - digit 1 or 0) are found from the code pattern 241 of the application code 130 (see FIG. 16), three are narrowed down (selected) by arithmetic intensity analysis. That is, as shown by reference symbol u in FIG. 17, the offload patterns "1000", "0010", and "0001" of three for loops are narrowed down from the four for loops.

[0301] <Example of "unfolding" executed together with code splitting during OpenCL conversion> When transferring data from an FPGA to a CPU, the loop statement described on the CPU program side [k = 0; k < 10; k++]{ } In this case, at the top of this loop statement, \pragma unroll is instructed. That is, \pragma unroll for(k = 0; k < 10; k++){ } is described.

[0302] When unroll is instructed in a grammar suitable for tools such as Intel's and Xilinx's (registered trademark) tools like \pragma unroll, in the above expansion example, it can be expanded to i = 0, i = 1, i = 2 and executed in a pipeline. For this reason, although the amount of resources used will be 10 times, it may become faster. Also, the number of expansions by unroll can be specified not as the total number of loop iterations but as 5 expansions, etc. In that case, every 2 loop iterations will be expanded into 5. The above concludes the explanation of the "expansion" example.

[0303] Next, the PLD processing pattern creation unit 215 further narrows down the loop statements with high arithmetic intensity that have been shortlisted as offload candidates, using the amount of resources. That is, the PLD processing pattern creation unit 215 calculates the amount of resources, and the PLD processing pattern creation unit 215 analyzes the resource efficiency (= arithmetic intensity / amount of resources during FPGA processing, or arithmetic intensity × number of loop iterations / amount of resources during FPGA processing) among the offload candidates of the loop statements with high arithmetic intensity, and extracts the loop statements with high resource efficiency.

[0304] In the symbol v in Fig. 17, the PLD processing pattern creation unit 215 compiles (<pre-compiles>) the OpenCL for offloading the narrowed-down loop statements.

[0305] <Specific Example of "High Arithmetic Intensity, OpenCL Conversion" (Part 2)> As shown by the symbol u in FIG. 17, out of the four offload patterns "1000", "0100", "0010", and "0001" narrowed down by the arithmetic intensity analysis, three offload patterns "1000", "0010", and "0001" are narrowed down by the above resource efficiency analysis. The above describes "high arithmetic intensity, OpenCL conversion" shown in step C of FIG. 17.

[0306] For the "loop statement with high resource efficiency" shown in step D of FIG. 17, the performance measurement unit 118 measures the performance of the compiled program ("first performance measurement"). Then, the PLD processing pattern creation unit 215 lists the loop statements that have been performance-measured and have higher performance than the CPU. Similarly hereinafter, the resource amount is calculated, offload OpenCL compilation is performed, and the performance of the compiled program is measured.

[0307] <Specific Example of "High Arithmetic Intensity, OpenCL Conversion" (Part 3)> As shown by the symbol w in FIG. 17, the first measurement is performed for the three offload patterns "1000", "0010", and "0001". Assuming that the performance of two of the three measurements, "1000" and "0010", is high, the second measurement is performed for the combination of "1000" and "0010".

[0308] In the symbol x of FIG. 17, the execution file creation unit 119 compiles OpenCL for offloading the narrowed-down loop statement (<this compilation>).

[0309] "Actual measurement of combination patterns" shown in step E of FIG. 17 means measuring the candidate loop statements alone and then the verification pattern measurements in combination.

[0310] <Specific Example of "High Arithmetic Intensity, OpenCL Conversion" (Part 4)> As shown by the reference sign y in FIG. 17, the combination of "1000" and "0010", namely "1010", is measured for the second time. After two measurements, as a result, "0010" with the highest speed among the first measurement and the second measurement is selected. In such a case, "0010" becomes the final solution. Here, there may be cases where the combination pattern cannot be measured due to resource quantity limitations. In this case, it is also possible to skip the combination and simply select the one with the highest speed from the individual results.

[0311] In the reference sign z in FIG. 17, the performance measurement unit 118 selects (<select>) "0010" with a good highest speed among the first measurement and the second measurement.

[0312] As described above, "0010" of the "OpenCL final solution" shown in the procedure F in FIG. 17 (refer to the reference sign aa in FIG. 17) is selected.

[0313] <Deployment (Placement)> In the PLD processing pattern with the highest processing performance of the OpenCL final solution, it is redeployed to the production environment and provided to the user.

[0314] [Implementation Example] An implementation example will be described. For the FPGA, Intel PAC with Intel Arria10 GX FPGA, etc. can be used. For FPGA processing, Intel Acceleration Stack (Intel FPGA SDK for OpenCL, Quartus Prime Version), etc. can be used. The Intel FPGA SDK for OpenCL is a high-level synthesis tool (HLS) that interprets #pragma, etc. for Intel in addition to standard OpenCL. In the implementation example, the OpenCL code describing the kernel processed by the FPGA and the host program processed by the CPU is interpreted, information such as the amount of resources is output, wiring work of the FPGA is performed, etc., so that it can operate on the FPGA. To enable operation on the actual FPGA, even a small program of about 100 lines takes about 3 hours. However, when the amount of resources exceeds the limit, an error occurs earlier. Also, in the case of OpenCL code that cannot be processed by the FPGA, an error is output after several hours.

[0315] In the implementation example, when there is a request to use a C / C++ application, first, the code of the C / C++ application is analyzed to find for loops and grasp the program structure such as variable data used within the for loops. For syntax analysis, syntax analysis libraries such as LLVM / Clang can be used.

[0316] In the implementation example, next, in order to obtain an expectation of whether there is an FPGA offloading effect for each loop statement, an arithmetic intensity analysis tool is executed to obtain an index of arithmetic intensity determined by the number of calculations, the number of accesses, etc. For arithmetic intensity analysis, frameworks such as the ROSE framework can be used. Only the loop statements with the top arithmetic intensity are targeted. Next, a profiling tool such as gcov is used to obtain the number of loop iterations for each loop. The loop statements with the top a in arithmetic intensity × number of loop iterations are narrowed down as candidates.

[0317] In the implementation example, next, for each individual loop statement with high arithmetic intensity, OpenCL code for FPGA offloading is generated. The OpenCL code is divided with the corresponding loop statement as the FPGA kernel and the rest as the CPU host program. When making it the FPGA kernel code, as a technique for speeding up, the loop statement expansion process may be performed by a certain number b. The loop statement expansion process increases the amount of resources but is effective for speeding up. Therefore, the number of expansions is limited to a certain number b and performed within a range where the amount of resources does not become huge.

[0318] In the implementation example, next, for a pieces of OpenCL code, pre-compilation is performed using the Intel FPGA SDK for OpenCL to calculate the amount of resources such as Flip Flop and Look Up Table to be used. The amount of resources used is displayed as a percentage of the total amount of resources. Here, the resource efficiency of each loop statement is calculated from the arithmetic intensity and the amount of resources, or from the arithmetic intensity, the number of loop iterations, and the amount of resources. For example, for a loop statement with an arithmetic intensity of 10 and a resource amount of 0.5, the resource efficiency is 10 / 0.5 = 20, and for a loop statement with an arithmetic intensity of 3 and a resource amount of 0.3, the resource efficiency is 3 / 0.3 = 10, and the former is higher. Also, the value multiplied by the number of loop iterations may be used as the resource efficiency. Select c loop statements with high resource efficiency in each loop statement.

[0319] In the implementation example, next, for c loop statements as candidates, patterns to be actually measured are created. For example, if the first and third loops have high resource efficiency, create each OpenCL pattern that offloads the first and offloads the third, compile it, and measure the performance. If the offload patterns of multiple single loop statements can achieve speedup (for example, if both the first and third can achieve speedup), create the OpenCL pattern in that combination, compile it, and measure the performance (for example, the pattern that offloads both the first and third).

[0320] Note that when creating combinations of single loops, the amount of resources used also combines. Therefore, if it does not fall within the upper limit value, that combination pattern is not created. When d patterns including combinations are created, perform performance measurement on the server equipped with the FPGA in the verification environment. For performance measurement, perform the sample processing specified in the application to be accelerated. For example, for a Fourier transform application, perform performance measurement using the conversion processing on sample data as a benchmark. In the implementation example, finally, select the fast patterns of the multiple measurement patterns as solutions.

[0321] In the second embodiment, the same "resource amount determination and placement determination" as described in the first embodiment is also executed (description omitted).

[0322] [Evaluation] Explain the evaluation. In the [FPGA Automatic Offloading of Loop Statements] of the second embodiment, evaluation can be performed in the same manner as in the [GPU Automatic Offloading of Loop Statements] of the first embodiment.

[0323] [Evaluation Target] The evaluation target is MRI-Q for MRI (Magnetic Resonance Imaging) image processing in the [FPGA Automatic Offloading of Loop Statements] of the second embodiment. MRI-Q calculates the matrix Q representing the scanner configuration used in the three-dimensional MRI reconstruction algorithm in non-Cartesian space. MRI-Q is written in C language, performs three-dimensional MRI image processing during performance measurement, and measures the processing time with data of the Large (maximum) size of 64×64×64. CPU processing uses C language, and FPGA processing is based on OpenCL.

[0324] [Evaluation Method] Input the code of the target application, and try loop statement offloading recognized by Clang or the like for the target GPU or FPGA to determine the offloading pattern. At this time, measure the processing time and power consumption. For the final offloading pattern, obtain the time change of the power consumption and confirm the power reduction compared to the case where all processing is performed by the CPU. In the [FPGA Automatic Offloading of Loop Statements] of the second embodiment, GA is not performed, and the measurement pattern is narrowed down to four patterns using arithmetic intensity or the like. Offloading Target Loop Statement: MRI-Q 16 Pattern Fitness: The lower the processing time, the higher the evaluation value and the higher the fitness. Also in MRI-Q of the second embodiment, in the same form as FIG. 12 described above, the cost and response time can be improved compared to the simple arrangement with priority on cost or response time.

[0325] [Hardware Configuration] The offloading server according to the first and second embodiments is realized by a computer 900 which is a physical device having a configuration as shown in FIG. 18, for example. FIG. 18 is a hardware configuration diagram showing an example of a computer that realizes the functions of the offload servers 1 and 1A. The computer 900 includes a CPU 901, a RAM 902, a ROM 903, an HDD 904, an accelerator 905, an input / output interface (I / F) 906, a media interface (I / F) 907, and a communication interface (I / F: Interface) 908.

[0326] The accelerator 905 is an accelerator (device) that processes at high speed at least one of the data from the communication I / F 908 and the data from the RAM 902. For example, the accelerator 905 is an accelerator for the various devices 151 in FIG. 2, the device 152 having a CPU-GPU, the device 153 having a CPU-FPGA, and the device 154 having a CPU. Note that, as the accelerator 905, a type (look-aside type) that returns the execution result to the CPU 901 or the RAM 902 after executing the processing from the CPU 901 or the RAM 902 may be used. On the other hand, as the accelerator 905, a type (in-line type) that enters between the communication I / F 908 and the CPU 901 or the RAM 902 and performs processing may be used.

[0327] The accelerator 905 is connected to an external device 915 via the communication I / F 908. The input / output I / F 906 is connected to an input / output device 916. The media I / F 907 reads and writes data from and to a recording medium 917.

[0328] The CPU 901 operates based on a program stored in the ROM 903 or the HDD 904, and executes the program (also called an application or an app for short) read into the RAM 902, thereby controlling each processing unit of the offload servers 1 and 1A shown in FIGS. 1 and 13. And this program can also be distributed via a communication line or recorded on a recording medium 917 such as a CD-ROM and distributed. The ROM 903 stores a boot program executed by the CPU 901 when the computer 900 is started up, programs dependent on the hardware of the computer 900, and the like.

[0329] The CPU 901 controls an input / output device 916 including an input unit such as a mouse and a keyboard, and an output unit such as a display and a printer, via the input / output I / F 906. The CPU 901 acquires data from the input / output device 916 via the input / output I / F 906, and outputs the generated data to the input / output device 916. Note that, together with the CPU 901, a GPU (Graphics Processing Unit) or the like may be used as a processor.

[0330] The HDD 904 stores programs executed by the CPU 901, data used by the programs, and the like. The communication I / F 908 receives data from other devices via a communication network (for example, NW (Network)) and outputs it to the CPU 901, and transmits data generated by the CPU 901 to other devices via the communication network.

[0331] The media I / F 907 reads a program or data stored in the recording medium 917 and outputs it to the CPU 901 via the RAM 902. The CPU 901 loads a program related to the target process from the recording medium 917 onto the RAM 902 via the media I / F 907, and executes the loaded program. The recording medium 917 is an optical recording medium such as a DVD (Digital Versatile Disc) or a PD (Phase change rewritable Disk), a magneto-optical recording medium such as an MO (Magneto Optical disk), a magnetic recording medium, a conductor memory tape medium, or a semiconductor memory or the like.

[0332] For example, when the computer 900 functions as the offload servers 1 and 1A according to the first and second embodiments, the CPU 901 of the computer 900 realizes the functions of the offload servers 1 and 1A by executing the program loaded on the RAM 902. Further, the HDD 904 stores the data in the RAM 902. The CPU 901 reads and executes the program related to the target process from the recording medium 912. Additionally, the CPU 901 may read the program related to the target process from another device via the communication network.

[0333] [Effect] As described above, the offload server 1 (see FIG. 1) according to the first embodiment is an offload server that offloads specific processing of an application program to an accelerator, and includes an application code analysis unit 112 that analyzes the source code of the application program, and analyzes the reference relationship of variables used in the loop statements of the application program, and for data that can be transferred outside the loop, performs data transfer designation using an explicit designation line that explicitly designates data transfer outside the loop. A data transfer designation unit 113, a parallel processing designation unit 114 that specifies loop statements of the application program and compiles them by designating parallel processing designation statements in the accelerator for each specified loop statement, and for loop statements that result in compilation errors, excludes them from the offload target, and for loop statements that do not result in compilation errors, creates a parallel processing pattern that designates whether to perform parallel processing or not. A parallel processing pattern creation unit 117, a performance measurement unit 118 that compiles the application program with the parallel processing pattern and arranges it in an accelerator verification device, and executes performance measurement processing when offloading to the accelerator, and the converted application program is stored in a cloud server, a carrier edge server, or a user edge server on the network according to the conditions of cost or response time specified by the user. When arranging, based on a linear programming equation with the cost of devices and links, the upper limit of computing resources, and the upper limit of bandwidth as constraints, and the cost of computing resources or response time as the objective function, calculates and sets the placement location of the application program. An arrangement setting unit 170 is provided.

[0334] By doing so, for an application automatically offloaded to an accelerator such as a GPU or FPGA, change the price conditions, response time conditions, number of application arrangements, etc. requested by the user, and calculate the appropriate arrangement to be placed in any of the cloud server, carrier edge server, or user edge server on the network. As a result, it is possible to realize an optimal arrangement that meets the requirements of the cost of computing resources or response time and conforms to the user's desires by satisfying the converted application.

[0335] The offload server 1A (see FIG. 13) according to the second embodiment is an offload server that offloads specific processing of an application program to a PLD, and includes an application code analysis unit 112 that analyzes the source code of the application program, a PLD process specification unit 213 that identifies loop statements of the application program and creates and compiles, for each identified loop statement, a plurality of offload process patterns in which pipeline processing and parallel processing in the PLD are specified by OpenCL, an arithmetic intensity calculation unit 214 that calculates the arithmetic intensity of the loop statements of the application program, a PLD process pattern creation unit 215 that narrows down loop statements with an arithmetic intensity higher than a predetermined threshold as offload candidates based on the arithmetic intensity calculated by the arithmetic intensity calculation unit 214 and creates a PLD process pattern, a performance measurement unit 118 that compiles the application program with the created PLD process pattern, arranges it in an accelerator verification device, and executes processing for performance measurement when offloaded to the PLD, and an arrangement setting unit 170 that calculates and sets the arrangement location of the application program based on a linear programming formula with the cost of devices and links, the upper limit of calculation resources, and the upper limit of bandwidth as constraint conditions and the cost of calculation resources or the response time as an objective function when arranging the converted application program on any one of a cloud server, a carrier edge server, and a user edge server on the network according to conditions of cost or response time specified by the user.

[0336] By doing so, after narrowing down the patterns to be actually performance - measured and then placing them in the verification environment, compiling, and performing performance measurement on the actual PLD (e.g., FPGA), the number of times of performance measurement can be reduced. As a result, in the automatic offloading to the PLD, the automatic offloading of the loop statements of the application can be performed at high speed. Then, for the converted application, by changing the price conditions, response - time conditions, number of application placements, etc. requested by the user, the optimal placement for placement on any one of the cloud server, carrier - edge server, and user - edge server on the network is calculated. Thereby, for the converted application, it is possible to realize the optimal placement that meets the requirements of the calculation resource cost or response time and conforms to the user's desires.

[0337] In the offloading servers 1, 1A according to the first and second embodiments, the placement setting unit 170 is characterized in that when an application program is placed on the server, it calculates a placement that minimizes the cost of the calculation resource or a placement that minimizes the response time.

[0338] By doing so, the converted application can be optimally placed while meeting the requirements of the calculation resource cost or response time.

[0339] In the offloading servers 1, 1A according to the first and second embodiments, the placement setting unit 170 is characterized in that when an application program is placed on the server, it calculates a placement that minimizes the cost of the calculation resource according to the linear programming formula shown in [Equation 1].

[0340] By doing so, for example, in the case of a requirement that placement is necessary within a certain amount per month, the minimization of the response time in Equation (1) of [Equation 1] becomes the objective function, and one of the constraint conditions is the maximum amount of the calculation resource cost in Equation (2) of [Equation 1]. Therefore, the converted application can be optimally placed while meeting the requirement of the calculation resource cost.

[0341] In the offload servers 1 and 1A according to the first and second embodiments, when an application program is arranged in the server, the arrangement setting unit 170 calculates an arrangement that minimizes the response time according to the linear programming formula shown in [Equation 2].

[0342] By doing so, for example, in the case of a requirement that the response time of the application needs to be within a certain number of seconds, the minimization of the cost of Equation (5) in [Equation 2] becomes the objective function, and that the response time of Equation (6) in [Equation 2] is within a certain number of seconds becomes one of the constraint conditions. Therefore, the converted application can be optimally arranged to meet the user's response time requirement.

[0343] The present invention uses an offload program to cause a computer to function as the offload server.

[0344] By doing so, each function of the offload servers 1 and 1A can be realized using a general computer.

[0345] Also, among the processes described in each of the above embodiments, all or part of the processes described as being automatically performed can be performed manually, or all or part of the processes described as being performed manually can be automatically performed by a known method. In addition, the processing procedures, control procedures, specific names, and information including various data and parameters shown in the above documents and drawings can be arbitrarily changed unless otherwise specified. Also, each component of each device shown in the drawings is conceptually functional and does not necessarily need to be physically configured as shown in the drawings. That is, the specific form of the distribution and integration of each device is not limited to that shown in the drawings, and all or part of it can be functionally or physically distributed and integrated in any unit according to various loads and usage situations.

[0346] Further, some or all of the above-described configurations, functions, processing units, processing means, etc. may be realized in hardware by designing them, for example, in an integrated circuit. Also, the above-described configurations, functions, etc. may be realized by software for a processor to interpret and execute a program for realizing each function. Information such as a program, table, file, etc. for realizing each function can be held in a memory, a recording device such as a hard disk, SSD (Solid State Drive), or a recording medium such as an IC (Integrated Circuit) card, SD (Secure Digital) card, or optical disk.

[0347] Also, in this embodiment, in order to be able to find a solution to the combinatorial optimization problem during a limited optimization period, a genetic algorithm (GA) method is used, but any optimization method may be used. For example, local search, Dynamic Programming, or a combination of these may be used.

[0348] Also, in this embodiment, an OpenACC compiler for C / C++ is used, but any compiler that can offload GPU processing may be used. For example, Java lambda (registered trademark) GPU processing or IBM Java 9 SDK (registered trademark) may be used. Note that the parallel processing specification statement depends on these development environments. For example, in Java (registered trademark), parallel processing can be described in lambda form since Java 8. IBM (registered trademark) provides a JIT compiler that offloads parallel processing in lambda form to a GPU. In Java, similar offloading is possible by tuning whether to make loop processing in lambda form or not using GA with these.

[0349] Also, in this embodiment, a for statement is exemplified as a repetition statement (loop statement), but while statements and do-while statements other than the for statement are also included. However, a for statement for specifying loop continuation conditions and the like is more suitable.

Description of Reference Numerals

[0350] 1,1A Offload Server 11,21 Control Unit 12 Input / Output Unit 13 Memory Unit 14 Verification Machine (Accelerator Verification Device) 111 Application Code Designation Unit 112 Application Code Analysis Unit 113 Data Transfer Designation Unit 114 Parallel Processing Designation Unit 114a,213a Offload Range Extraction Unit 114b,213b Intermediate Language File Output Unit 115 Resource Ratio Determination Unit 116 Resource Amount Setting Unit 117 Parallel Processing Pattern Creation Unit 118 Performance Measurement Unit 118a Binary File Placement Unit 119 Execution File Creation Unit 120 Production Environment Placement Unit 121 Performance Measurement Test Extraction Execution Unit 122 User Provision Unit 125 Application Code 131 Test Case DB 132 Facility Resource DB 133 Intermediate Language File 151 Various Devices 152 Device with CPU-GPU 153 Device with CPU-FPGA 154 Device with CPU 170 Placement Setting Unit 213 PLD Processing Designation Unit 214 Arithmetic Intensity Calculation Unit 215 PLD Processing Pattern Creation Unit 905 Accelerator

Claims

1. An offloading server that offloads specific processing of an application program to an accelerator, comprising: an application code analysis unit that analyzes the source code of the application program; a data transfer specification unit that analyzes the reference relationship of variables used in the loop statements of the application program, and performs data transfer specification using an explicit specification line that explicitly specifies data transfer outside the loop for data that can be transferred outside the loop; a parallel processing specification unit that identifies the loop statements of the application program, and for each identified loop statement, specifies and compiles a parallel processing specification statement in the accelerator; a parallel processing pattern creation unit that creates a parallel processing pattern for specifying whether to perform parallel processing or not for loop statements that do not result in compilation errors, while excluding loop statements that result in compilation errors from the offloading target; a performance measurement unit that compiles the application program of the parallel processing pattern, arranges it in an accelerator verification device, and executes performance measurement processing when offloading to the accelerator; an arrangement setting unit that calculates and sets the arrangement location of the application program based on a linear programming equation with the cost of devices and links, the upper limit of computing resources, and the upper limit of bandwidth as constraint conditions, and the cost of computing resources or response time as the objective function when arranging the converted application program on any of a cloud server, a carrier edge server, or a user edge server on the network according to the conditions of cost or response time specified by the user; The arrangement setting unit calculates an arrangement that minimizes the response time according to the following linear programming equation when arranging the application program on the server. 【Number 1】 An offloading server characterized by the above.

2. An offloading server that offloads specific processing of an application program to an accelerator, comprising: an application code analysis unit that analyzes the source code of the application program; a data transfer specification unit that analyzes the reference relationship of variables used in the loop statements of the application program, and performs data transfer specification using an explicit specification line that explicitly specifies data transfer outside the loop for data that can be transferred outside the loop; A parallel processing specification unit that identifies loop statements in the application program and, for each identified loop statement, specifies a parallel processing specification statement in the accelerator and compiles it; A parallel processing pattern creation unit that creates a parallel processing pattern that excludes loop statements that result in compilation errors from offloading targets and specifies whether to perform parallel processing or not for loop statements that do not result in compilation errors; A performance measurement unit that compiles the application program of the parallel processing pattern, places it in an accelerator verification device, and executes a performance measurement process when offloading to the accelerator; When placing the converted application program on any of a cloud server, a carrier edge server, or a user edge server on the network according to the cost or response time conditions specified by the user, a device and link cost, a calculation resource upper limit, and a bandwidth upper limit are used as constraint conditions, and a placement setting unit that calculates and sets the placement location of the application program based on a linear programming formula with the cost of the calculation resource or the response time as the objective function; When placing the application program on the server, the placement setting unit calculates the placement that minimizes the cost of the calculation resources according to the linear programming formula shown in the following equation 【Number 2】 An offloading server characterized by the above.

3. An offloading server that offloads the specific processing of the application program to a PLD (Programmable Logic Device), An application code analysis unit that analyzes the source code of the application program; A PLD processing specification unit that identifies loop statements in the application program and, for each identified loop statement, creates and compiles a plurality of offloading processing patterns specified by pipeline processing and parallel processing in the PLD using OpenCL; An arithmetic intensity calculation unit that calculates the arithmetic intensity of the loop statements in the application program; Based on the arithmetic intensity calculated by the arithmetic intensity calculation unit, a PLD processing pattern creation unit that narrows down loop statements with an arithmetic intensity higher than a predetermined threshold as offloading candidates and creates a PLD processing pattern; Compile the application program of the generated PLD processing pattern, place it in the accelerator verification device, and execute the performance measurement process when offloading to the PLD. A performance measurement unit; When the converted application program is placed on any of a cloud server, a carrier edge server, or a user edge server on the network according to the cost or response time conditions specified by the user, the costs of devices and links, the upper limit of computing resources, and the upper limit of bandwidth are used as constraints, and based on a linear programming formula with the cost of computing resources or the response time as the objective function, calculate and set the placement location of the application program. A placement setting unit; When the placement setting unit places the application program on the server, it calculates the placement that minimizes the response time according to the linear programming formula shown in the following formula [Number 3] An offloading server characterized by the above.

4. An offloading server that offloads specific processing of an application program to a PLD (Programmable Logic Device), An application code analysis unit that analyzes the source code of the application program; Identify the loop statements of the application program, and for each identified loop statement, create and compile a plurality of offloading processing patterns specified in OpenCL for pipeline processing and parallel processing in the PLD. A PLD processing specification unit; An arithmetic intensity calculation unit that calculates the arithmetic intensity of the loop statements of the application program; Based on the arithmetic intensity calculated by the arithmetic intensity calculation unit, narrow down the loop statements with an arithmetic intensity higher than a predetermined threshold as offloading candidates, and create a PLD processing pattern. A PLD processing pattern creation unit; Compile the application program of the generated PLD processing pattern, place it in the accelerator verification device, and execute the performance measurement process when offloading to the PLD. A performance measurement unit; When placing the converted application program on any of a cloud server, a carrier edge server, or a user edge server on the network according to the cost or response time conditions specified by the user, the cost of devices and links, the upper limit of computing resources, and the upper limit of bandwidth are used as constraints, and based on a linear programming formula with the cost of computing resources or the response time as the objective function, an arrangement setting unit that calculates and sets the placement location of the application program is provided. When the arrangement setting unit places an application program on a server, it calculates an arrangement that minimizes the cost of computing resources according to a linear programming formula shown in the following formula. 【Number 4】 An offload server characterized by the above.

5. An offload control method for an offload server that offloads specific processing of an application program to an accelerator, The offload server is Analyzing the source code of the application program, Analyzing the reference relationship of variables used in the loop statements of the application program, and for data that can be transferred outside the loop, performing data transfer designation using an explicit designation line that explicitly designates data transfer outside the loop. Identifying the loop statements of the application program, and for each identified loop statement, specifying a parallel processing designation statement in the accelerator and compiling it. For loop statements that result in compilation errors, excluding them from the offload target, and creating a parallel processing pattern that specifies whether to perform parallel processing or not for loop statements that do not result in compilation errors. Compiling the application program of the parallel processing pattern, placing it on an accelerator verification device, and executing a performance measurement process when offloading to the accelerator. When placing the converted application program on any of a cloud server, a carrier edge server, or a user edge server on the network according to the cost or response time conditions specified by the user, the cost of devices and links, the upper limit of computing resources, and the upper limit of bandwidth are used as constraints, and based on a linear programming formula with the cost of computing resources or the response time as the objective function, calculating and setting the placement location of the application program is executed. In the step of calculating and setting the placement location of the application program, when the application program is placed on the server, the placement that minimizes the response time is calculated according to the linear programming formula shown in the following formula [Number 5] An offloading control method characterized by this.

6. An offloading control method for an offloading server that offloads specific processing of an application program to an accelerator, The offloading server,[[]]END]] analyzing the source code of the application program; analyzing the reference relationship of variables used in the loop statements of the application program, and for data that can be transferred outside the loop, performing data transfer designation using an explicit designation line that explicitly designates data transfer outside the loop; identifying the loop statements of the application program, and for each identified loop statement, specifying a parallel processing designation statement in the accelerator and compiling it; for loop statements that result in compilation errors, excluding them from the offloading target, and creating a parallel processing pattern for specifying whether to perform parallel processing or not for loop statements that do not result in compilation errors; compiling the application program of the parallel processing pattern, placing it in an accelerator verification device, and executing a performance measurement process when offloading to the accelerator; When placing the converted application program on any of a cloud server, a carrier edge server, or a user edge server on the network according to the conditions of cost or response time specified by the user, using the costs of devices and links, the upper limit of computing resources, and the upper limit of bandwidth as constraints, and calculating and setting the placement location of the application program based on a linear programming formula with the cost of computing resources or the response time as the objective function; In the step of calculating and setting the placement location of the application program, when the application program is placed on the server, the placement that minimizes the cost of computing resources is calculated according to the linear programming formula shown in the following formula 【Number 6】 An offloading control method characterized by this.

7. An offloading control method for an offloading server that offloads specific processing of an application program to a PLD (Programmable Logic Device), wherein the offloading server analyzes the source code of the application program; identifies loop statements of the application program, and for each identified loop statement, creates and compiles the pipeline processing and parallel processing in the PLD according to a plurality of offloading processing patterns specified by OpenCL; calculates the arithmetic intensity of the loop statements of the application program; narrow down loop statements with an arithmetic intensity higher than a predetermined threshold as offloading candidates based on the calculated arithmetic intensity, and create a PLD processing pattern; compiles the application program of the created PLD processing pattern, arranges it in an accelerator verification device, and executes performance measurement processing when offloading to the PLD; when arranging the converted application program on any of a cloud server, a carrier edge server, and a user edge server on a network according to the conditions of cost or response time specified by the user, calculate and set the placement location of the application program based on a linear programming formula with the cost of devices and links, the upper limit of computing resources, and the upper limit of bandwidth as constraint conditions, and the cost of computing resources or the response time as the objective function, and execute; In the step of calculating and setting the placement location of the application program, when arranging the application program on a server, calculate the placement that minimizes the response time according to the linear programming formula shown in the following formula 【Number 7】 The offloading control method is characterized by the above.

8. An offloading control method for an offloading server that offloads specific processing of an application program to a PLD (Programmable Logic Device), wherein the offloading server analyzes the source code of the application program; identifies loop statements of the application program, and for each identified loop statement, creates and compiles the pipeline processing and parallel processing in the PLD according to a plurality of offloading processing patterns specified by OpenCL; A step of calculating the arithmetic intensity of the loop statement of the application program; Based on the calculated arithmetic intensity, narrowing down the loop statements with the arithmetic intensity higher than a predetermined threshold as offloading candidates, and creating a PLD processing pattern; Compiling the application program of the created PLD processing pattern, arranging it in an accelerator verification device, and executing a performance measurement process when offloading to the PLD; When arranging the converted application program on any of a cloud server, a carrier edge server, and a user edge server on the network according to the cost or response time conditions specified by the user, using the cost of devices and links, the upper limit of computing resources, and the upper limit of bandwidth as constraint conditions, and calculating and setting the placement location of the application program based on a linear programming formula with the cost of computing resources or the response time as the objective function; In the step of calculating and setting the placement location of the application program, when arranging the application program on the server, calculating the placement that minimizes the cost of computing resources according to the following linear programming formula 【Number 8】 An offloading control method characterized by the above.

9. An offloading program for causing a computer to function as the offloading server according to any one of Claims 1 to 4.

Images