cancel
Showing results for 
Search instead for 
Did you mean: 

Archives Discussions

settle
Challenger

AMD APP KernelAnalyzer and cl_amd_vec3

I'm using AMD APP KernelAnalyzer 1.12.1288 with CAL 12.4 and attempting to use vec3 data types in my kernels.

Source Code:

#pragma OPENCL EXTENSION cl_amd_vec3 : enable

__kernel void u3d(uint3 n, uint3 pitch, float h, __global float *u)

{

    uint3 global_offset = (uint3)(get_global_offset(0), get_global_id(1), get_global_id(2));

    uint3 global_id = (uint3)(get_global_id(0), get_global_id(1), get_global_id(2));

    uint l = dot(convert_float3(global_id - global_offset), convert_float3(pitch));

    float3 x = -1 + convert_float3(global_id - global_offset) * h;

    u = dot(sin(M_PI_F * x), 1);

}

If I include the pragma line I get:

OpenCL Compile Error: clBuildProgram failed (CL_BUILD_PROGRAM_FAILURE).

error: can't enable all OpenCL extensions or unrecognized OpenCL extension

  #pragma OPENCL EXTENSION cl_amd_vec3 : enable

                                                                       ^

1 error detected.

If I exclude the pragma line I get:

OpenCL Compile Error: clBuildProgram failed (CL_BUILD_PROGRAM_FAILURE).

calclCompile failedError: Creating kernel u3d failed!

Can anyone tell me what I may be doing wrong or if this is a problem with the KernelAnalyzer?  Thanks!

0 Likes
1 Solution
rouellet
Staff

This appears to be a bug in the OpenCL support.

When I add these pragmas to one of the APP SDK samples I see the same errors.

I used the BoxFilter example, because it has code to report compilation errors.

I will report open a bug report.

c:\Users\rouellet.AMD\Documents\AMD APP\samples\opencl\bin\x86_64>BoxFilter.exe

Running SAT version..

Platform 0 : Advanced Micro Devices, Inc.

Platform found : Advanced Micro Devices, Inc.

Selected Platform Vendor : Advanced Micro Devices, Inc.

Device 0 : Capeverde Device ID is 00000000001DF650

                        BUILD LOG

************************************************

"C:\Users\rouellet.AMD\AppData\Local\Temp\OCLE085.tmp.cl", line 1: error:

          can't enable all OpenCL extensions or unrecognized OpenCL extension

  #pragma OPENCL EXTENSION cl_amd_vec3 : enable

                                         ^

"C:\Users\rouellet.AMD\AppData\Local\Temp\OCLE085.tmp.cl", line 2: warning:

          ignore unrecognized OpenCL extension

  #pragma OPENCL EXTENSION cl_khr_fp64 : enable

                                               ^

1 error detected in the compilation of "C:\Users\rouellet.AMD\AppData\Local\Temp\OCLE085.tmp.cl".

Internal error: clc compiler invocation failed.

************************************************

Error: clBuildProgram failed. Error code : CL_BUILD_PROGRAM_FAILURE

Location : SDKCommon.cpp:1614

Error: sampleCommon::buildOpenCLProgram() failed

Location : BoxFilterSAT.cpp:338

View solution in original post

0 Likes
14 Replies
binying
Challenger

One question, by using the command clinfo, you see that the extension cl_amd_vec3 is supported on your GPUs?

0 Likes

My devices all support cl_amd_vec3, but I'm specifically talking about running my kernels through the AMD APP KernelAnalyzer so that shouldn't be an issue anyways.

0 Likes

well, I am trying to repeat it on my desktop first...

0 Likes

I'm also using AMD APP KernelAnalyzer 1.12.1288 with CAL 12.4. Any of the following extension causes compiling error although they are in the list of supported extensions. So I think this is a tool issue. Hopefully,  this is helpful for you.

//------------------------------------------------------------------------

#pragma OPENCL EXTENSION cl_khr_int64_base_atomics : enable

#pragma OPENCL EXTENSION cl_khr_int64_extended_atomics : enable

#pragma OPENCL EXTENSION cl_khr_gl_sharing : enable

#pragma OPENCL EXTENSION cl_ext_device_fission : enable

#pragma OPENCL EXTENSION cl_amd_device_attribute_query : enable

#pragma OPENCL EXTENSION cl_amd_vec3 : enable

#pragma OPENCL EXTENSION cl_khr_d3d10_sharing : enable

__kernel void u3d(uint4 n, uint4 pitch, float h, __global float *u)

{

          //uint3 global_offset = (uint3)(get_global_offset(0), get_global_id(1), get_global_id(2));

          //uint3 global_id = (uint3)(get_global_id(0), get_global_id(1), get_global_id(2));

          // uint l = dot(convert_float3(global_id - global_offset), convert_float3(pitch));

          //float3 x = -1 + convert_float3(global_id - global_offset) * h;

          // u = dot(sin(M_PI_F * x), 1);

}

0 Likes

you need enable only that extensions in kernel code which have effect on kernel language. so gl_sharing, d3d10_sharing device_stribute_query, device_fission are redudant in code.

0 Likes

Aha~.  Thank you, nou!

0 Likes

Why has this post been marked "assumed answered" and there is no way to remove that status?  Nothing has been answered about why this is happening or how to fix it, only confirmed that the problem exists.

0 Likes

You can choose one response and mark it as Correct Answer, and then unmark it, the status will become Not answered.

0 Likes
rouellet
Staff

This appears to be a bug in the OpenCL support.

When I add these pragmas to one of the APP SDK samples I see the same errors.

I used the BoxFilter example, because it has code to report compilation errors.

I will report open a bug report.

c:\Users\rouellet.AMD\Documents\AMD APP\samples\opencl\bin\x86_64>BoxFilter.exe

Running SAT version..

Platform 0 : Advanced Micro Devices, Inc.

Platform found : Advanced Micro Devices, Inc.

Selected Platform Vendor : Advanced Micro Devices, Inc.

Device 0 : Capeverde Device ID is 00000000001DF650

                        BUILD LOG

************************************************

"C:\Users\rouellet.AMD\AppData\Local\Temp\OCLE085.tmp.cl", line 1: error:

          can't enable all OpenCL extensions or unrecognized OpenCL extension

  #pragma OPENCL EXTENSION cl_amd_vec3 : enable

                                         ^

"C:\Users\rouellet.AMD\AppData\Local\Temp\OCLE085.tmp.cl", line 2: warning:

          ignore unrecognized OpenCL extension

  #pragma OPENCL EXTENSION cl_khr_fp64 : enable

                                               ^

1 error detected in the compilation of "C:\Users\rouellet.AMD\AppData\Local\Temp\OCLE085.tmp.cl".

Internal error: clc compiler invocation failed.

************************************************

Error: clBuildProgram failed. Error code : CL_BUILD_PROGRAM_FAILURE

Location : SDKCommon.cpp:1614

Error: sampleCommon::buildOpenCLProgram() failed

Location : BoxFilterSAT.cpp:338

0 Likes

Hi Roland,

Thank you for your follow-up.  One other thing related to cl_amd_vec3 I just noticed is that the usertype.dat file included in AMD APP SDK 2.7 does not include syntax highlighting for those types (and some convert_* functions seem to be missing).  I've attached an updated file that I think fills in the blanks.

Cheers!

Sean

0 Likes

I have created another bug report.

http://www.amd.com/report will redirect you to a form with a URL like:

http://www.amdsurveys.com/se.ashx?s=5A1E27D20B2F3EBB

I think this is the official bug reporting mechanism.

Thanks,

R.

You can also contact AMD support either by email (http://emailcustomercare.amd.com/) or by phone (http://support.amd.com/us/contacts/Pages/global-technical-support.aspx) to report bugs.

The email form is nice in that it lets you send attachments.

A fix has been checked into the development branch for the vec3 & fp64 problems.

I cannot say for sure when users will see the fix, but the process usually takes several months.

0 Likes
settle
Challenger

Just a quick follow-up about the original issue with cl_amd_vec3 not being recognized.  If you query a device using clGetDeviceInfo for CL_DEVICE_OPENCL_C_VERSION and it returns OpenCL C 1.1 or 1.2 (and probably anything thereafter), then vec3's are already supported as per those specifications, so no extensions required.   So maybe the cl_amd_vec3 extension is only an issue for OpenCL C 1.0 devices.

0 Likes