<html xmlns:v="urn:schemas-microsoft-com:vml" xmlns:o="urn:schemas-microsoft-com:office:office" xmlns:w="urn:schemas-microsoft-com:office:word" xmlns:m="http://schemas.microsoft.com/office/2004/12/omml" xmlns="http://www.w3.org/TR/REC-html40">
<head>
<meta http-equiv="Content-Type" content="text/html; charset=us-ascii">
<meta name="Generator" content="Microsoft Word 14 (filtered medium)">
<style><!--
/* Font Definitions */
@font-face
{font-family:Calibri;
panose-1:2 15 5 2 2 2 4 3 2 4;}
@font-face
{font-family:"Lucida Console";
panose-1:2 11 6 9 4 5 4 2 2 4;}
/* Style Definitions */
p.MsoNormal, li.MsoNormal, div.MsoNormal
{margin:0in;
margin-bottom:.0001pt;
font-size:11.0pt;
font-family:"Calibri","sans-serif";}
a:link, span.MsoHyperlink
{mso-style-priority:99;
color:blue;
text-decoration:underline;}
a:visited, span.MsoHyperlinkFollowed
{mso-style-priority:99;
color:purple;
text-decoration:underline;}
span.EmailStyle17
{mso-style-type:personal-compose;
font-family:"Calibri","sans-serif";
color:windowtext;}
.MsoChpDefault
{mso-style-type:export-only;
font-family:"Calibri","sans-serif";}
@page WordSection1
{size:8.5in 11.0in;
margin:1.0in 1.0in 1.0in 1.0in;}
div.WordSection1
{page:WordSection1;}
--></style><!--[if gte mso 9]><xml>
<o:shapedefaults v:ext="edit" spidmax="1026" />
</xml><![endif]--><!--[if gte mso 9]><xml>
<o:shapelayout v:ext="edit">
<o:idmap v:ext="edit" data="1" />
</o:shapelayout></xml><![endif]-->
</head>
<body lang="EN-US" link="blue" vlink="purple">
<div class="WordSection1">
<p class="MsoNormal">I am not sure if this is the appropriate list on which to ask this question, if not hopefully someone can suggest an alternative.<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">Under Linux, I am attempting to perform an offline compile of an OpenCL kernel example using Clang, and then load that binary using the clCreateProgramWithBinary() function.<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">Unfortunately, while clover is loading the binary, I end up getting a segmentation fault:<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal" style="margin-left:.5in;text-autospace:none"><span style="font-size:9.0pt;font-family:"Lucida Console"">Program received signal SIGSEGV, Segmentation fault.<o:p></o:p></span></p>
<p class="MsoNormal" style="margin-left:.5in;text-autospace:none"><span style="font-size:9.0pt;font-family:"Lucida Console"">proc (v=..., is=...) at core/module.cpp:50<o:p></o:p></span></p>
<p class="MsoNormal" style="margin-left:.5in;text-autospace:none"><span style="font-size:9.0pt;font-family:"Lucida Console"">50 T x;<o:p></o:p></span></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">I have pasted the source code I am using below, for both the kernel and the host code.<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">I am compiling both with the following commands:<o:p></o:p></p>
<p class="MsoNormal">clang -target r600-unknown-unknown -x cl -S -emit-llvm -mcpu=r600 kernel.cl -o kernel.clbin
<o:p></o:p></p>
<p class="MsoNormal">clang -g -L/usr/local/lib -lOpenCL offline_host.c -o offline_host<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">I have LLVM/Clang 3.4RC3 installed and Mesa 10.0.1.<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">If anyone has suggestions, or can point me to the appropriate mailing list or documentation, I’d appreciate it.<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">Thanks!<o:p></o:p></p>
<p class="MsoNormal">-Al<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">Source code for “kernel.cl”<o:p></o:p></p>
<p class="MsoNormal">====================<o:p></o:p></p>
<p class="MsoNormal">__kernel void vecAdd(__global float* a)<o:p></o:p></p>
<p class="MsoNormal">{<o:p></o:p></p>
<p class="MsoNormal"> int gid = get_global_id(0);<o:p></o:p></p>
<p class="MsoNormal"> a[gid] += a[gid];<o:p></o:p></p>
<p class="MsoNormal">}<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">Source code for “offline_host.c”<o:p></o:p></p>
<p class="MsoNormal">==========================<o:p></o:p></p>
<p class="MsoNormal">#include <stdio.h><o:p></o:p></p>
<p class="MsoNormal">#include <stdlib.h><o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">#ifdef __APPLE__<o:p></o:p></p>
<p class="MsoNormal">#include <OpenCL/opencl.h><o:p></o:p></p>
<p class="MsoNormal">#else<o:p></o:p></p>
<p class="MsoNormal">#include <CL/cl.h><o:p></o:p></p>
<p class="MsoNormal">#endif<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">#define MEM_SIZE (128)<o:p></o:p></p>
<p class="MsoNormal">#define MAX_BINARY_SIZE (0x100000)<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal">int main()<o:p></o:p></p>
<p class="MsoNormal">{<o:p></o:p></p>
<p class="MsoNormal"> cl_platform_id platform_id = NULL;<o:p></o:p></p>
<p class="MsoNormal"> cl_device_id device_id = NULL;<o:p></o:p></p>
<p class="MsoNormal"> cl_context context = NULL;<o:p></o:p></p>
<p class="MsoNormal"> cl_command_queue command_queue = NULL;<o:p></o:p></p>
<p class="MsoNormal"> cl_mem memobj = NULL;<o:p></o:p></p>
<p class="MsoNormal"> cl_program program = NULL;<o:p></o:p></p>
<p class="MsoNormal"> cl_kernel kernel = NULL;<o:p></o:p></p>
<p class="MsoNormal"> cl_uint ret_num_devices;<o:p></o:p></p>
<p class="MsoNormal"> cl_uint ret_num_platforms;<o:p></o:p></p>
<p class="MsoNormal"> cl_int ret;<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> float mem[MEM_SIZE];<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> FILE *fp;<o:p></o:p></p>
<p class="MsoNormal"> char fileName[] = "kernel.clbin";<o:p></o:p></p>
<p class="MsoNormal"> size_t binary_size;<o:p></o:p></p>
<p class="MsoNormal"> char *binary_buf;<o:p></o:p></p>
<p class="MsoNormal"> cl_int binary_status;<o:p></o:p></p>
<p class="MsoNormal"> cl_int i;<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Load kernel binary */<o:p></o:p></p>
<p class="MsoNormal"> fp = fopen(fileName, "r");<o:p></o:p></p>
<p class="MsoNormal"> if (!fp) {<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> }<o:p></o:p></p>
<p class="MsoNormal"> binary_buf = (char *)malloc(MAX_BINARY_SIZE);<o:p></o:p></p>
<p class="MsoNormal"> binary_size = fread(binary_buf, 1, MAX_BINARY_SIZE, fp);<o:p></o:p></p>
<p class="MsoNormal"> fclose(fp);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Initialize input data */<o:p></o:p></p>
<p class="MsoNormal"> for (i = 0; i < MEM_SIZE; i++) {<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> }<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Get platform/device information */<o:p></o:p></p>
<p class="MsoNormal"> ret = clGetPlatformIDs(1, &platform_id, &ret_num_platforms);<o:p></o:p></p>
<p class="MsoNormal"> ret = clGetDeviceIDs(platform_id, CL_DEVICE_TYPE_GPU, 1, &device_id, &ret_num_devices);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Create OpenCL context*/<o:p></o:p></p>
<p class="MsoNormal"> context = clCreateContext(NULL, 1, &device_id, NULL, NULL, &ret);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Create command queue */<o:p></o:p></p>
<p class="MsoNormal"> command_queue = clCreateCommandQueue(context, device_id, 0, &ret);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Create memory buffer */<o:p></o:p></p>
<p class="MsoNormal"> memobj = clCreateBuffer(context, CL_MEM_READ_WRITE, MEM_SIZE * sizeof(float), NULL, &ret);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Transfer data over to the memory buffer */<o:p></o:p></p>
<p class="MsoNormal"> ret = clEnqueueWriteBuffer(command_queue, memobj, CL_TRUE, 0, MEM_SIZE * sizeof(float), mem, 0, NULL, NULL);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Create kernel program from the kernel binary */<o:p></o:p></p>
<p class="MsoNormal"> program = clCreateProgramWithBinary(context, 1, &device_id, (const size_t *)&binary_size,<o:p></o:p></p>
<p class="MsoNormal"> (const unsigned char **)&binary_buf, &binary_status, &ret);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Create OpenCL kernel */<o:p></o:p></p>
<p class="MsoNormal"> kernel = clCreateKernel(program, "vecAdd", &ret);<o:p></o:p></p>
<p class="MsoNormal"> printf("err:%d\n", ret);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Set OpenCL kernel arguments */<o:p></o:p></p>
<p class="MsoNormal"> ret = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *)&memobj);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> size_t global_work_size[3] = {MEM_SIZE, 0, 0};<o:p></o:p></p>
<p class="MsoNormal"> size_t local_work_size[3] = {MEM_SIZE, 0, 0};<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Execute OpenCL kernel */<o:p></o:p></p>
<p class="MsoNormal"> ret = clEnqueueNDRangeKernel(command_queue, kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Copy result from the memory buffer */<o:p></o:p></p>
<p class="MsoNormal"> ret = clEnqueueReadBuffer(command_queue, memobj, CL_TRUE, 0, MEM_SIZE * sizeof(float), mem, 0, NULL, NULL);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Display results */<o:p></o:p></p>
<p class="MsoNormal"> for (i=0; i < MEM_SIZE; i++) {<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> }<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> /* Finalization */<o:p></o:p></p>
<p class="MsoNormal"> ret = clFlush(command_queue);<o:p></o:p></p>
<p class="MsoNormal"> ret = clFinish(command_queue);<o:p></o:p></p>
<p class="MsoNormal"> ret = clReleaseKernel(kernel);<o:p></o:p></p>
<p class="MsoNormal"> ret = clReleaseProgram(program);<o:p></o:p></p>
<p class="MsoNormal"> ret = clReleaseMemObject(memobj);<o:p></o:p></p>
<p class="MsoNormal"> ret = clReleaseCommandQueue(command_queue);<o:p></o:p></p>
<p class="MsoNormal"> ret = clReleaseContext(context);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> free(binary_buf);<o:p></o:p></p>
<p class="MsoNormal"><o:p> </o:p></p>
<p class="MsoNormal"> return 0;<o:p></o:p></p>
<p class="MsoNormal">}<o:p></o:p></p>
<p class="MsoNormal"><i><span style="font-size:10.0pt"><o:p> </o:p></span></i></p>
</div>
</body>
</html>