source: CLRX/CLRadeonExtender/trunk/samples/CLUtils.cpp @ 3450

Last change on this file since 3450 was 3450, checked in by matszpk, 2 years ago

CLRadeonExtender: Add new library: CLRXCLHelper that facilitate creating binary/program for OpenCL. Use CLRXCLHelper in samples.

File size: 9.5 KB
Line 
1/*
2 *  CLRadeonExtender - Unofficial OpenCL Radeon Extensions Library
3 *  Copyright (C) 2014-2017 Mateusz Szpakowski
4 *
5 *  This library is free software; you can redistribute it and/or
6 *  modify it under the terms of the GNU Lesser General Public
7 *  License as published by the Free Software Foundation; either
8 *  version 2.1 of the License, or (at your option) any later version.
9 *
10 *  This library is distributed in the hope that it will be useful,
11 *  but WITHOUT ANY WARRANTY; without even the implied warranty of
12 *  MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the GNU
13 *  Lesser General Public License for more details.
14 *
15 *  You should have received a copy of the GNU Lesser General Public
16 *  License along with this library; if not, write to the Free Software
17 *  Foundation, Inc., 51 Franklin Street, Fifth Floor, Boston, MA  02110-1301  USA
18 */
19
20#define CL_USE_DEPRECATED_OPENCL_1_2_APIS
21#define CL_USE_DEPRECATED_OPENCL_2_0_APIS
22
23#include <algorithm>
24#include <iostream>
25#include <fstream>
26#include <vector>
27#include <cstring>
28#include <memory>
29#include <CL/cl.h>
30#include <CLRX/amdasm/Assembler.h>
31#include <CLRX/clhelper/CLHelper.h>
32#include "CLUtils.h"
33
34using namespace CLRX;
35
36/* parse args from command line and handle options:
37 * print help, list OpenCL devices, get choosen device and choosen OpenCL standard
38 * return true if sample should exit */
39bool CLFacade::parseArgs(const char* progName, const char* usagePart, int argc,
40                  const char** argv, cl_uint& deviceIndex, cxuint& useCL)
41{
42    if (argc >= 2 && ::strcmp(argv[1], "-?")==0)
43    {
44        std::cout << "Usage: " << progName << " [DEVICE_INDEX[cl1|old|cl2]] " <<
45                usagePart << "\n" "Print device list: " << progName << " -L" << "\n"
46                "Print help: " << progName << " -?\n"
47                "'cl2' after DEVICE_INDEX enables AMD OpenCL 2.0 mode\n"
48                "'cl1' or 'old' after DEVICE_INDEX force old AMD OpenCL 1.2 mode"
49                << std::endl;
50        return true;
51    }
52   
53    cl_int error = CL_SUCCESS;
54    cl_platform_id choosenPlatform = chooseCLPlatformForCLRX();
55   
56    if (argc >= 2 && ::strcmp(argv[1], "-L")==0)
57    {
58        // list devices, before it get GPU devices
59        cl_uint devicesNum;
60        std::unique_ptr<cl_device_id[]> devices;
61        error = clGetDeviceIDs(choosenPlatform, CL_DEVICE_TYPE_GPU, 0,
62                               nullptr, &devicesNum);
63        if (error != CL_SUCCESS)
64            throw CLError(error, "clGetDeviceIDs");
65       
66        devices.reset(new cl_device_id[devicesNum]);
67        error = clGetDeviceIDs(choosenPlatform, CL_DEVICE_TYPE_GPU,
68                        devicesNum, devices.get(), nullptr);
69        if (error != CL_SUCCESS)
70            throw CLError(error, "clGetDeviceIDs");
71       
72        for (cl_uint i = 0; i < devicesNum; i++)
73        {
74            cl_device_id device = devices[i];
75            // get device and print that
76            size_t deviceNameSize;
77            std::unique_ptr<char[]> deviceName;
78            error = clGetDeviceInfo(device, CL_DEVICE_NAME, 0, nullptr, &deviceNameSize);
79            if (error != CL_SUCCESS)
80                throw CLError(error, "clGetDeviceInfoName");
81            deviceName.reset(new char[deviceNameSize]);
82            error = clGetDeviceInfo(device, CL_DEVICE_NAME, deviceNameSize,
83                                     deviceName.get(), nullptr);
84            if (error != CL_SUCCESS)
85                throw CLError(error, "clGetDeviceInfoName");
86            std::cout << "Device: " << i << " - " << deviceName.get() << "\n";
87        }
88        std::cout.flush();
89        return true;
90    }
91    else if (argc >= 2)
92    {
93        const char* end;
94        useCL = 0;
95        deviceIndex = cstrtovCStyle<cl_uint>(argv[1], nullptr, end);
96        if (strcasecmp(end, "cl2")==0)
97            useCL = 2;
98        else if (strcasecmp(end, "cl1")==0 || strcasecmp(end, "old")==0)
99            useCL = 1;
100    }
101    return false;
102}
103
104static const char* binaryFormatNamesTbl[] =
105{
106    "AMD OpenCL 1.2", "GalliumCompute", "Raw code", "AMD OpenCL 2.0"
107};
108
109CLFacade::CLFacade(cl_uint deviceIndex, const char* sourceCode, const char* kernelNames,
110            cxuint useCL)
111{
112try
113{
114    context = nullptr;
115    device = nullptr;
116    queue = nullptr;
117    program = nullptr;
118   
119    cl_int error = CL_SUCCESS;
120    const cl_platform_id choosenPlatform = chooseCLPlatformForCLRX();
121   
122    // find device
123    cl_uint devicesNum;
124    std::unique_ptr<cl_device_id[]> devices;
125    error = clGetDeviceIDs(choosenPlatform, CL_DEVICE_TYPE_GPU, 0, nullptr, &devicesNum);
126    if (error != CL_SUCCESS)
127        throw CLError(error, "clGetDeviceIDs");
128   
129    if (deviceIndex >= devicesNum)
130        throw CLError(0, "DeviceIndexOutOfRange");
131       
132    devices.reset(new cl_device_id[devicesNum]);
133    error = clGetDeviceIDs(choosenPlatform, CL_DEVICE_TYPE_GPU,
134                    devicesNum, devices.get(), nullptr);
135    if (error != CL_SUCCESS)
136        throw CLError(error, "clGetDeviceIDs");
137   
138    device = devices[deviceIndex];
139   
140    const CLAsmSetup asmSetup = assemblerSetupForCLDevice(device, useCL==1 ?
141            CLHELPER_USEAMDLEGACY : useCL==2 ? CLHELPER_USEAMDCL2 : 0);
142   
143    // get workGroupSize and Compute Units of device
144    error = clGetDeviceInfo(device, CL_DEVICE_MAX_WORK_GROUP_SIZE, sizeof(size_t),
145                             &maxWorkGroupSize, nullptr);
146    if (error != CL_SUCCESS)
147        throw CLError(error, "clGetDeviceMaxWorkGroupSize");
148    error = clGetDeviceInfo(device, CL_DEVICE_MAX_COMPUTE_UNITS, sizeof(cl_uint),
149                             &computeUnits, nullptr);
150    if (error != CL_SUCCESS)
151        throw CLError(error, "clGetDeviceMaxComputeUnits");
152   
153    /// create context
154    context = clCreateContext(nullptr, 1, &device, nullptr, nullptr, &error);
155    if (context==nullptr)
156        throw CLError(error, "clCreateContext");
157   
158    // get device and print that
159    size_t deviceNameSize;
160    std::unique_ptr<char[]> deviceName;
161    error = clGetDeviceInfo(device, CL_DEVICE_NAME, 0, nullptr, &deviceNameSize);
162    if (error != CL_SUCCESS)
163        throw CLError(error, "clGetDeviceInfoName");
164    deviceName.reset(new char[deviceNameSize]);
165    error = clGetDeviceInfo(device, CL_DEVICE_NAME, deviceNameSize,
166                             deviceName.get(), nullptr);
167    if (error != CL_SUCCESS)
168        throw CLError(error, "clGetDeviceInfoName");
169    std::cout << "Device: " << deviceIndex << " - " << deviceName.get() << std::endl;
170   
171    std::cout << "BinaryFormat: " <<
172        binaryFormatNamesTbl[cxuint(asmSetup.binaryFormat)] << "\n"
173        "Bitness: " << (asmSetup.is64Bit ? 64U : 32U) << std::endl;
174   
175    /// create command queue
176    queue = clCreateCommandQueue(context, device, 0, &error);
177    if (queue==nullptr)
178        throw CLError(error, "clCreateCommandQueue");
179   
180    CString buildLog;
181    try
182    { program = createProgramForCLDevice(context, device,
183                    asmSetup, sourceCode, 0, &buildLog); }
184    // function throw exception when fail
185    catch(const CLError& error)
186    {
187        std::cerr << "BuildLog:\n" << buildLog << std::endl;
188        throw;
189    }
190   
191    if (kernelNames!=nullptr)
192    try
193    {
194        for (const char* kn = kernelNames; *kn!=0;)
195        {
196            const char* knameStart = kn;
197            while (*kn!=0 && *kn!=' ') kn++;
198            std::string kernelName(knameStart, kn);
199            cl_kernel kernel = clCreateKernel(program, kernelName.c_str(), &error);
200            if (error != CL_SUCCESS)
201                throw CLError(error, "clCreateKernel");
202            kernels.push_back(kernel);
203            while (*kn==' ') kn++; // skip spaces
204        }
205    }
206    catch(...)
207    {
208        for (cl_kernel kernel: kernels)
209            clReleaseKernel(kernel);
210        throw;
211    }
212}
213catch(...)
214{
215    if (program!=nullptr)
216        clReleaseProgram(program);
217    if (queue!=nullptr)
218        clReleaseCommandQueue(queue);
219    if (context!=nullptr)
220        clReleaseContext(context);
221    throw;
222}
223}
224
225CLFacade::~CLFacade()
226{
227    for (cl_mem memObj: memObjects)
228        clReleaseMemObject(memObj);
229    for (cl_kernel kernel: kernels)
230        clReleaseKernel(kernel);
231    if (program!=nullptr)
232        clReleaseProgram(program);
233    if (queue!=nullptr)
234        clReleaseCommandQueue(queue);
235    if (context!=nullptr)
236        clReleaseContext(context);
237}
238
239// get work group size and work group size multiple from kernel
240void CLFacade::getKernelInfo(cl_kernel kernel, size_t& workGroupSize,
241               size_t& workGroupSizeMultiple)
242{
243    cl_int error = clGetKernelWorkGroupInfo(kernel, device, CL_KERNEL_WORK_GROUP_SIZE,
244                 sizeof(size_t), &workGroupSize, nullptr);
245    if (error != CL_SUCCESS)
246        throw CLError(error, "clGetKernelWorkGroupSize");
247    error = clGetKernelWorkGroupInfo(kernel, device,
248                 CL_KERNEL_PREFERRED_WORK_GROUP_SIZE_MULTIPLE,
249                 sizeof(size_t), &workGroupSizeMultiple, nullptr);
250    if (error != CL_SUCCESS || workGroupSizeMultiple == 1)
251        workGroupSizeMultiple = 64; // fix for GalliumCompute
252}
253
254void CLFacade::callNDRangeKernel(cl_kernel kernel, cl_uint workDim, const size_t* offset,
255               const size_t* workSize, const size_t* localSize)
256{
257    cl_event event;
258    cl_int error = clEnqueueNDRangeKernel(queue, kernel, workDim, offset, workSize,
259                                   localSize, 0, nullptr, &event);
260    if (error != CL_SUCCESS)
261        throw CLError(error, "clEnqueueNDRangeKernel");
262    error = clWaitForEvents(1, &event); // waiting for finish kernel
263    if (error != CL_SUCCESS)
264    {
265        clReleaseEvent(event);
266        throw CLError(error, "clWaitForEvents");
267    }
268    clReleaseEvent(event);
269}
Note: See TracBrowser for help on using the repository browser.