111308Santhony.gutierrez@amd.com/* 211308Santhony.gutierrez@amd.com * Copyright (c) 2015 Advanced Micro Devices, Inc. 311308Santhony.gutierrez@amd.com * All rights reserved. 411308Santhony.gutierrez@amd.com * 511308Santhony.gutierrez@amd.com * For use for simulation and test purposes only 611308Santhony.gutierrez@amd.com * 711308Santhony.gutierrez@amd.com * Redistribution and use in source and binary forms, with or without 811308Santhony.gutierrez@amd.com * modification, are permitted provided that the following conditions are met: 911308Santhony.gutierrez@amd.com * 1011308Santhony.gutierrez@amd.com * 1. Redistributions of source code must retain the above copyright notice, 1111308Santhony.gutierrez@amd.com * this list of conditions and the following disclaimer. 1211308Santhony.gutierrez@amd.com * 1311308Santhony.gutierrez@amd.com * 2. Redistributions in binary form must reproduce the above copyright notice, 1411308Santhony.gutierrez@amd.com * this list of conditions and the following disclaimer in the documentation 1511308Santhony.gutierrez@amd.com * and/or other materials provided with the distribution. 1611308Santhony.gutierrez@amd.com * 1711308Santhony.gutierrez@amd.com * 3. Neither the name of the copyright holder nor the names of its contributors 1811308Santhony.gutierrez@amd.com * may be used to endorse or promote products derived from this software 1911308Santhony.gutierrez@amd.com * without specific prior written permission. 2011308Santhony.gutierrez@amd.com * 2111308Santhony.gutierrez@amd.com * THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" 2211308Santhony.gutierrez@amd.com * AND ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE 2311308Santhony.gutierrez@amd.com * IMPLIED WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE 2411308Santhony.gutierrez@amd.com * ARE DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE 2511308Santhony.gutierrez@amd.com * LIABLE FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR 2611308Santhony.gutierrez@amd.com * CONSEQUENTIAL DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF 2711308Santhony.gutierrez@amd.com * SUBSTITUTE GOODS OR SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS 2811308Santhony.gutierrez@amd.com * INTERRUPTION) HOWEVER CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN 2911308Santhony.gutierrez@amd.com * CONTRACT, STRICT LIABILITY, OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) 3011308Santhony.gutierrez@amd.com * ARISING IN ANY WAY OUT OF THE USE OF THIS SOFTWARE, EVEN IF ADVISED OF THE 3111308Santhony.gutierrez@amd.com * POSSIBILITY OF SUCH DAMAGE. 3211308Santhony.gutierrez@amd.com * 3311308Santhony.gutierrez@amd.com * Author: Marc Orr, Brad Beckmann 3411308Santhony.gutierrez@amd.com */ 3511308Santhony.gutierrez@amd.com 3611308Santhony.gutierrez@amd.com#include <CL/cl.h> 3711308Santhony.gutierrez@amd.com#include <malloc.h> 3811308Santhony.gutierrez@amd.com 3911308Santhony.gutierrez@amd.com#include <cstdio> 4011308Santhony.gutierrez@amd.com#include <cstring> 4111308Santhony.gutierrez@amd.com#include <fstream> 4211308Santhony.gutierrez@amd.com#include <string> 4311308Santhony.gutierrez@amd.com 4411615Sdavid.j.hashe@gmail.com#ifdef KVM_SWITCH 4511615Sdavid.j.hashe@gmail.com#include "m5op.h" 4611615Sdavid.j.hashe@gmail.com 4711615Sdavid.j.hashe@gmail.comvoid *m5_mem = (void*)0xffffc90000000000; 4811615Sdavid.j.hashe@gmail.com#endif 4911615Sdavid.j.hashe@gmail.com 5011308Santhony.gutierrez@amd.com#define SUCCESS 0 5111308Santhony.gutierrez@amd.com#define FAILURE 1 5211308Santhony.gutierrez@amd.com 5311308Santhony.gutierrez@amd.com// OpenCL datastructures 5411308Santhony.gutierrez@amd.comcl_context context; 5511308Santhony.gutierrez@amd.comcl_device_id *devices; 5611308Santhony.gutierrez@amd.comcl_command_queue commandQueue; 5711308Santhony.gutierrez@amd.comcl_program program; 5811308Santhony.gutierrez@amd.comcl_kernel readKernel; 5911308Santhony.gutierrez@amd.com 6011308Santhony.gutierrez@amd.com// Application datastructures 6111308Santhony.gutierrez@amd.comconst int CACHE_LINE_SIZE = 64; 6211308Santhony.gutierrez@amd.comsize_t grid_size = 512; 6311308Santhony.gutierrez@amd.comsize_t work_group_size = 256; 6411308Santhony.gutierrez@amd.com 6511308Santhony.gutierrez@amd.com// arguments 6611308Santhony.gutierrez@amd.comconst int code_size = 5; 6711308Santhony.gutierrez@amd.comconst char *code = "hello"; 6811308Santhony.gutierrez@amd.comint *keys; 6911308Santhony.gutierrez@amd.comchar *msg; 7011308Santhony.gutierrez@amd.comint chars_decoded = 0; 7111308Santhony.gutierrez@amd.com 7211308Santhony.gutierrez@amd.com/* 7311308Santhony.gutierrez@amd.com Setup data structures for application/algorithm 7411308Santhony.gutierrez@amd.com*/ 7511308Santhony.gutierrez@amd.comint 7611308Santhony.gutierrez@amd.comsetupDataStructs() 7711308Santhony.gutierrez@amd.com{ 7811308Santhony.gutierrez@amd.com msg = (char *)memalign(CACHE_LINE_SIZE, (grid_size + 1) * sizeof(char)); 7911321Ssteve.reinhardt@amd.com if (msg == NULL) { 8011308Santhony.gutierrez@amd.com printf("%s:%d: error: %s\n", __FILE__, __LINE__, 8111308Santhony.gutierrez@amd.com "could not allocate host buffers\n"); 8211308Santhony.gutierrez@amd.com exit(-1); 8311308Santhony.gutierrez@amd.com } 8411308Santhony.gutierrez@amd.com msg[grid_size] = '\0'; 8511308Santhony.gutierrez@amd.com 8611308Santhony.gutierrez@amd.com keys = (int *)memalign(CACHE_LINE_SIZE, code_size * sizeof(int)); 8711308Santhony.gutierrez@amd.com keys[0] = 23; 8811308Santhony.gutierrez@amd.com keys[1] = 0; 8911308Santhony.gutierrez@amd.com keys[2] = 0; 9011308Santhony.gutierrez@amd.com keys[3] = 0; 9111308Santhony.gutierrez@amd.com keys[4] = 0; 9211308Santhony.gutierrez@amd.com 9311308Santhony.gutierrez@amd.com return SUCCESS; 9411308Santhony.gutierrez@amd.com} 9511308Santhony.gutierrez@amd.com 9611308Santhony.gutierrez@amd.com/* Setup OpenCL data structures */ 9711308Santhony.gutierrez@amd.comint 9811308Santhony.gutierrez@amd.comsetupOpenCL() 9911308Santhony.gutierrez@amd.com{ 10011308Santhony.gutierrez@amd.com cl_int status = 0; 10111308Santhony.gutierrez@amd.com size_t deviceListSize; 10211308Santhony.gutierrez@amd.com 10311308Santhony.gutierrez@amd.com // 1. Get platform 10411308Santhony.gutierrez@amd.com cl_uint numPlatforms; 10511308Santhony.gutierrez@amd.com cl_platform_id platform = NULL; 10611308Santhony.gutierrez@amd.com status = clGetPlatformIDs(0, NULL, &numPlatforms); 10711308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 10811308Santhony.gutierrez@amd.com printf("Error: Getting Platforms. (clGetPlatformsIDs)\n"); 10911308Santhony.gutierrez@amd.com return FAILURE; 11011308Santhony.gutierrez@amd.com } 11111308Santhony.gutierrez@amd.com 11211308Santhony.gutierrez@amd.com if (numPlatforms > 0) { 11311308Santhony.gutierrez@amd.com cl_platform_id *platforms = new cl_platform_id[numPlatforms]; 11411308Santhony.gutierrez@amd.com status = clGetPlatformIDs(numPlatforms, platforms, NULL); 11511308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 11611308Santhony.gutierrez@amd.com printf("Error: Getting Platform Ids. (clGetPlatformsIDs)\n"); 11711308Santhony.gutierrez@amd.com return FAILURE; 11811308Santhony.gutierrez@amd.com } 11911308Santhony.gutierrez@amd.com for (int i = 0; i < numPlatforms; ++i) { 12011308Santhony.gutierrez@amd.com char pbuff[100]; 12111308Santhony.gutierrez@amd.com status = clGetPlatformInfo(platforms[i], CL_PLATFORM_VENDOR, 12211308Santhony.gutierrez@amd.com sizeof(pbuff), pbuff, NULL); 12311308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 12411308Santhony.gutierrez@amd.com printf("Error: Getting Platform Info.(clGetPlatformInfo)\n"); 12511308Santhony.gutierrez@amd.com return FAILURE; 12611308Santhony.gutierrez@amd.com } 12711308Santhony.gutierrez@amd.com platform = platforms[i]; 12811308Santhony.gutierrez@amd.com if (!strcmp(pbuff, "Advanced Micro Devices, Inc.")) { 12911308Santhony.gutierrez@amd.com break; 13011308Santhony.gutierrez@amd.com } 13111308Santhony.gutierrez@amd.com } 13211308Santhony.gutierrez@amd.com delete platforms; 13311308Santhony.gutierrez@amd.com } 13411308Santhony.gutierrez@amd.com 13511321Ssteve.reinhardt@amd.com if (NULL == platform) { 13611308Santhony.gutierrez@amd.com printf("NULL platform found so Exiting Application.\n"); 13711308Santhony.gutierrez@amd.com return FAILURE; 13811308Santhony.gutierrez@amd.com } 13911308Santhony.gutierrez@amd.com 14011308Santhony.gutierrez@amd.com // 2. create context from platform 14111308Santhony.gutierrez@amd.com cl_context_properties cps[3] = 14211308Santhony.gutierrez@amd.com {CL_CONTEXT_PLATFORM, (cl_context_properties)platform, 0}; 14311308Santhony.gutierrez@amd.com context = clCreateContextFromType(cps, CL_DEVICE_TYPE_GPU, NULL, NULL, 14411308Santhony.gutierrez@amd.com &status); 14511308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 14611308Santhony.gutierrez@amd.com printf("Error: Creating Context. (clCreateContextFromType)\n"); 14711308Santhony.gutierrez@amd.com return FAILURE; 14811308Santhony.gutierrez@amd.com } 14911308Santhony.gutierrez@amd.com 15011308Santhony.gutierrez@amd.com // 3. Get device info 15111308Santhony.gutierrez@amd.com // 3a. Get # of devices 15211308Santhony.gutierrez@amd.com status = clGetContextInfo(context, CL_CONTEXT_DEVICES, 0, NULL, 15311308Santhony.gutierrez@amd.com &deviceListSize); 15411308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 15511308Santhony.gutierrez@amd.com printf("Error: Getting Context Info (1st clGetContextInfo)\n"); 15611308Santhony.gutierrez@amd.com return FAILURE; 15711308Santhony.gutierrez@amd.com } 15811308Santhony.gutierrez@amd.com 15911308Santhony.gutierrez@amd.com // 3b. Get the device list data 16011308Santhony.gutierrez@amd.com devices = (cl_device_id *)malloc(deviceListSize); 16111308Santhony.gutierrez@amd.com if (devices == 0) { 16211308Santhony.gutierrez@amd.com printf("Error: No devices found.\n"); 16311308Santhony.gutierrez@amd.com return FAILURE; 16411308Santhony.gutierrez@amd.com } 16511308Santhony.gutierrez@amd.com status = clGetContextInfo(context, CL_CONTEXT_DEVICES, deviceListSize, 16611308Santhony.gutierrez@amd.com devices, NULL); 16711308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 16811308Santhony.gutierrez@amd.com printf("Error: Getting Context Info (2nd clGetContextInfo)\n"); 16911308Santhony.gutierrez@amd.com return FAILURE; 17011308Santhony.gutierrez@amd.com } 17111308Santhony.gutierrez@amd.com 17211308Santhony.gutierrez@amd.com // 4. Create command queue for device 17311308Santhony.gutierrez@amd.com commandQueue = clCreateCommandQueue(context, devices[0], 0, &status); 17411308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 17511308Santhony.gutierrez@amd.com printf("Creating Command Queue. (clCreateCommandQueue)\n"); 17611308Santhony.gutierrez@amd.com return FAILURE; 17711308Santhony.gutierrez@amd.com } 17811308Santhony.gutierrez@amd.com 17911308Santhony.gutierrez@amd.com const char *source = "dummy text"; 18011308Santhony.gutierrez@amd.com 18111308Santhony.gutierrez@amd.com size_t sourceSize[] = {strlen(source)}; 18211308Santhony.gutierrez@amd.com 18311308Santhony.gutierrez@amd.com // 5b. Register the kernel with the runtime 18411308Santhony.gutierrez@amd.com program = clCreateProgramWithSource(context, 1, &source, sourceSize, 18511308Santhony.gutierrez@amd.com &status); 18611308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 18711308Santhony.gutierrez@amd.com printf("Error: Loading kernel (clCreateProgramWithSource)\n"); 18811308Santhony.gutierrez@amd.com return FAILURE; 18911308Santhony.gutierrez@amd.com } 19011308Santhony.gutierrez@amd.com 19111308Santhony.gutierrez@amd.com status = clBuildProgram(program, 1, devices, NULL, NULL, NULL); 19211308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 19311308Santhony.gutierrez@amd.com printf("Error: Building kernel (clBuildProgram)\n"); 19411308Santhony.gutierrez@amd.com return FAILURE; 19511308Santhony.gutierrez@amd.com } 19611308Santhony.gutierrez@amd.com 19711308Santhony.gutierrez@amd.com readKernel = clCreateKernel(program, "read_kernel", &status); 19811308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 19911308Santhony.gutierrez@amd.com printf("Error: Creating readKernel from program. (clCreateKernel)\n"); 20011308Santhony.gutierrez@amd.com return FAILURE; 20111308Santhony.gutierrez@amd.com } 20211308Santhony.gutierrez@amd.com 20311308Santhony.gutierrez@amd.com return SUCCESS; 20411308Santhony.gutierrez@amd.com} 20511308Santhony.gutierrez@amd.com 20611308Santhony.gutierrez@amd.com 20711308Santhony.gutierrez@amd.com/* Run kernels */ 20811308Santhony.gutierrez@amd.comint 20911308Santhony.gutierrez@amd.comrunCLKernel(cl_kernel kernel) 21011308Santhony.gutierrez@amd.com{ 21111308Santhony.gutierrez@amd.com cl_int status; 21211308Santhony.gutierrez@amd.com cl_event event; 21311308Santhony.gutierrez@amd.com size_t globalThreads[1] = {grid_size}; 21411308Santhony.gutierrez@amd.com size_t localThreads[1] = {work_group_size}; 21511308Santhony.gutierrez@amd.com 21611308Santhony.gutierrez@amd.com // 1. Set arguments 21711308Santhony.gutierrez@amd.com // 1a. code size 21811308Santhony.gutierrez@amd.com size_t code_size = strlen(code); 21911308Santhony.gutierrez@amd.com status = clSetKernelArg(kernel, 0, sizeof(size_t), &code_size); 22011308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 22111308Santhony.gutierrez@amd.com printf("Error: Setting kernel argument. (code_size)\n"); 22211308Santhony.gutierrez@amd.com return FAILURE; 22311308Santhony.gutierrez@amd.com } 22411308Santhony.gutierrez@amd.com 22511308Santhony.gutierrez@amd.com // 1b. code 22611308Santhony.gutierrez@amd.com status = clSetKernelArg(kernel, 1, sizeof(char *), (void *)&code); 22711308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 22811308Santhony.gutierrez@amd.com printf("Error: Setting kernel argument. (code_in)\n"); 22911308Santhony.gutierrez@amd.com return FAILURE; 23011308Santhony.gutierrez@amd.com } 23111308Santhony.gutierrez@amd.com 23211308Santhony.gutierrez@amd.com // 1c. keys 23311308Santhony.gutierrez@amd.com printf("keys = %p, &keys = %p, keys[0] = %d\n", keys, &keys, keys[0]); 23411308Santhony.gutierrez@amd.com status = clSetKernelArg(kernel, 2, sizeof(int *), (void *)&keys); 23511308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 23611308Santhony.gutierrez@amd.com printf("Error: Setting kernel argument. (key_arr)\n"); 23711308Santhony.gutierrez@amd.com return FAILURE; 23811308Santhony.gutierrez@amd.com } 23911308Santhony.gutierrez@amd.com 24011308Santhony.gutierrez@amd.com // 1d. msg 24111308Santhony.gutierrez@amd.com status = clSetKernelArg(kernel, 3, sizeof(char *), (void *)&msg); 24211308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 24311308Santhony.gutierrez@amd.com printf("Error: Setting kernel argument. (memOut)\n"); 24411308Santhony.gutierrez@amd.com return FAILURE; 24511308Santhony.gutierrez@amd.com } 24611308Santhony.gutierrez@amd.com 24711308Santhony.gutierrez@amd.com // 1e. chars_decoded 24811308Santhony.gutierrez@amd.com int *chars_decoded_ptr = &chars_decoded; 24911308Santhony.gutierrez@amd.com status = clSetKernelArg(kernel, 4, sizeof(int *), 25011308Santhony.gutierrez@amd.com (void *)&chars_decoded_ptr); 25111308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 25211308Santhony.gutierrez@amd.com printf("Error: Setting kernel argument. (memOut)\n"); 25311308Santhony.gutierrez@amd.com return FAILURE; 25411308Santhony.gutierrez@amd.com } 25511308Santhony.gutierrez@amd.com 25611615Sdavid.j.hashe@gmail.com#ifdef KVM_SWITCH 25711615Sdavid.j.hashe@gmail.com m5_switchcpu(); 25811615Sdavid.j.hashe@gmail.com#endif 25911615Sdavid.j.hashe@gmail.com 26011308Santhony.gutierrez@amd.com // 2. Launch kernel 26111308Santhony.gutierrez@amd.com status = clEnqueueNDRangeKernel(commandQueue, kernel, 1, NULL, 26211308Santhony.gutierrez@amd.com globalThreads, localThreads, 0, NULL, 26311308Santhony.gutierrez@amd.com &event); 26411308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 26511308Santhony.gutierrez@amd.com printf("Error: Enqueue failed. (clEnqueueNDRangeKernel)\n"); 26611308Santhony.gutierrez@amd.com return FAILURE; 26711308Santhony.gutierrez@amd.com } 26811308Santhony.gutierrez@amd.com 26911308Santhony.gutierrez@amd.com // 3. Wait for the kernel 27011308Santhony.gutierrez@amd.com status = clWaitForEvents(1, &event); 27111308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 27211308Santhony.gutierrez@amd.com printf("Error: Waiting for kernel run to finish. (clWaitForEvents)\n"); 27311308Santhony.gutierrez@amd.com return FAILURE; 27411308Santhony.gutierrez@amd.com } 27511308Santhony.gutierrez@amd.com 27611308Santhony.gutierrez@amd.com // 4. Cleanup 27711308Santhony.gutierrez@amd.com status = clReleaseEvent(event); 27811308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 27911308Santhony.gutierrez@amd.com printf("Error: Release event object. (clReleaseEvent)\n"); 28011308Santhony.gutierrez@amd.com return FAILURE; 28111308Santhony.gutierrez@amd.com } 28211308Santhony.gutierrez@amd.com 28311308Santhony.gutierrez@amd.com return SUCCESS; 28411308Santhony.gutierrez@amd.com} 28511308Santhony.gutierrez@amd.com 28611308Santhony.gutierrez@amd.com 28711308Santhony.gutierrez@amd.com/* Release OpenCL resources (Context, Memory etc.) */ 28811308Santhony.gutierrez@amd.comint 28911308Santhony.gutierrez@amd.comcleanupCL() 29011308Santhony.gutierrez@amd.com{ 29111308Santhony.gutierrez@amd.com cl_int status; 29211308Santhony.gutierrez@amd.com status = clReleaseKernel(readKernel); 29311308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 29411308Santhony.gutierrez@amd.com printf("Error: In clReleaseKernel \n"); 29511308Santhony.gutierrez@amd.com return FAILURE; 29611308Santhony.gutierrez@amd.com } 29711308Santhony.gutierrez@amd.com status = clReleaseProgram(program); 29811308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 29911308Santhony.gutierrez@amd.com printf("Error: In clReleaseProgram\n"); 30011308Santhony.gutierrez@amd.com return FAILURE; 30111308Santhony.gutierrez@amd.com } 30211308Santhony.gutierrez@amd.com status = clReleaseCommandQueue(commandQueue); 30311308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 30411308Santhony.gutierrez@amd.com printf("Error: In clReleaseCommandQueue\n"); 30511308Santhony.gutierrez@amd.com return FAILURE; 30611308Santhony.gutierrez@amd.com } 30711308Santhony.gutierrez@amd.com status = clReleaseContext(context); 30811308Santhony.gutierrez@amd.com if (status != CL_SUCCESS) { 30911308Santhony.gutierrez@amd.com printf("Error: In clReleaseContext\n"); 31011308Santhony.gutierrez@amd.com return FAILURE; 31111308Santhony.gutierrez@amd.com } 31211308Santhony.gutierrez@amd.com 31311308Santhony.gutierrez@amd.com return SUCCESS; 31411308Santhony.gutierrez@amd.com} 31511308Santhony.gutierrez@amd.com 31611308Santhony.gutierrez@amd.comint 31711308Santhony.gutierrez@amd.commain(int argc, char * argv[]) 31811308Santhony.gutierrez@amd.com{ 31911308Santhony.gutierrez@amd.com // Initialize Host application 32011308Santhony.gutierrez@amd.com if (setupDataStructs() != SUCCESS) { 32111308Santhony.gutierrez@amd.com return FAILURE; 32211308Santhony.gutierrez@amd.com } 32311308Santhony.gutierrez@amd.com 32411308Santhony.gutierrez@amd.com // Initialize OpenCL resources 32511308Santhony.gutierrez@amd.com if (setupOpenCL() != SUCCESS) { 32611308Santhony.gutierrez@amd.com return FAILURE; 32711308Santhony.gutierrez@amd.com } 32811308Santhony.gutierrez@amd.com 32911308Santhony.gutierrez@amd.com // Run the CL program 33011308Santhony.gutierrez@amd.com if (runCLKernel(readKernel) != SUCCESS) { 33111308Santhony.gutierrez@amd.com return FAILURE; 33211308Santhony.gutierrez@amd.com } 33311308Santhony.gutierrez@amd.com printf("the gpu says:\n"); 33411308Santhony.gutierrez@amd.com printf("%s\n", msg); 33511308Santhony.gutierrez@amd.com 33611308Santhony.gutierrez@amd.com // Releases OpenCL resources 33711308Santhony.gutierrez@amd.com if (cleanupCL()!= SUCCESS) { 33811308Santhony.gutierrez@amd.com return FAILURE; 33911308Santhony.gutierrez@amd.com } 34011308Santhony.gutierrez@amd.com 34111308Santhony.gutierrez@amd.com return SUCCESS; 34211308Santhony.gutierrez@amd.com} 343