in geeks3d it was said HelloCL.exe was not working..
Let begin the investigation..
On Windows x64 and Linux (either 32 o amd64) already was working..
NOTE: (for using with Nvidia OpenCL SDK)
===============================
Some people at Nvidia forums don't know how to do, just use libraries (.libs) and in cl_platform.h use CL_API_CALL __stdcall (see AMD header).. Also if you want OpenCL OpenGL interop use cl_gl.h from AMD SDK or the current cl_gl.h at Khronos..
First I have created simple scripts in Windows to check all the samples for error!
Download them!
(only a.bat b.bat and test.bat are needed others are for fixes Nvidia OpenCL SDK post)
extract on AMD executables directory and from shell type test.bat
This is thanks to most samples supporting -e flag to verify results.. it prints passed or error!
Also i fyou want to get rid of all other stuff use -q..
My scripts call all the executables with these flags and redirect the console output stuff to a file..
Now test this with Nvidia OpenCL implementation..
log
===
Running AESEncryptDecrypt
Failed
Running BinarySearch
Passed!
Running BinomialOption
Passed!
Running BitonicSort
Passed!
Running BlackScholes
Passed!
Running DCT
Passed!
Running DwtHaar1D
Passed!
Running EigenValue
Passed!
Running FastWalshTransform
Passed!
Running FloydWarshall
Passed!
Running HelloCL
HelloCL!
Creating a context
Running Histogram
Error: clEnqueueNDRangeKernel failed. Error code : CL_OUT_OF_RESOURCES
Running Mandelbrot
Error: clEnqueueNDRangeKernel failed. Error code : CL_OUT_OF_RESOURCES
Running MatrixMultiplication
Passed!
Running MatrixTranspose
Passed!
Running MersenneTwister
Error: clBuildProgram failed. Error code : unknown error code
Running MonteCarloAsian
Error: clBuildProgram failed. Error code : unknown error code
Running NBody
Passed!
Running PrefixSum
Passed!
Running QuasiRandomSequence
Error: clBuildProgram failed. Error code : unknown error code
Running RadixSort
Failed
Running RecursiveGaussian
Passed!
Running Reduction
Passed!
Running ScanLargeArrays
Passed!
Running SimpleConvolution
Passed!
Running SobelFilter
Error: clBuildProgram failed. Error code : unknown error code
Running Template
Input:
0 1 2 3 4 5 6 7 8 9 ..
Error: Creating Context. (clCreateContextFromType)
Running TemplateC
Input:
0 1 2 3 4 5 6 7 8 9 ..
Error: Creating Context. (clCreateContextFromType)
Error: Setting kernel argument. (output)
Output:
3452816845 ..
Error: In clReleaseKernel
if you see that a lot work.. but HelloCL no..
There errors are due to four kinds of errors:
*Samples hardcoded to use CPU backend (Nvidia has not anyone..) (Error: Creating Context. (clCreateContextFromType))
*Kernel source (.cl) doesn't compile due to bugs in implementations or cases no defined in OpenCL spec (Error: clBuildProgram failed. Error code : unknown error code)
*Not enough resources on Nvidia GPUs or OpenCL implementation.. (clEnqueueNDRangeKernel failed. Error code : CL_OUT_OF_RESOURCES)
*Computation isn't correct (Failed! not Passed!)
of this kinds:
CPU:HelloCL,Template,TemplateC
Kernel source: HelloCL,MersenneTwister,MonteCarloAsian,QuasiRandomSequence,SobelFilter
Resources:Histogram,Mandelbrot
Results fail: AESEncryptDecrypt,RadixSort,SobelFilter
Now with trivial fixes you can get..
CPU errors fixed
Kernel source: errors fixed
Resources: erros fixed
Results fail: I have no fixed them..
For fixing CPU errors change clCreateDeviceformtupye (CL_DEVICE_TYPE_CPU to GPU)
Kernel source errors:
==============
All bugs related to converting int to float not working:
a. uint4 to float4
b. finding math functions passing ints
1.hellocl
Nvidia compiler returns
:5: error: a __kernel function cannot have varargs or stdargs
__kernel void
the kernel hasn't parameters:
__kernel void main()
this is a bug likely, note no problem for practical purposes, for Nvidia
fix putting:
__kernel void hello(uint width)
now you have to associate the parameter if not then you get at executing kernel (enqueueNdrangekernel) and error: -52
related to setting parameters so add
kernel.setArg(0, 0);
before executing and the example is fixed
2. and 3. montecarlo and mersene samples
This is likely a bug in Nvidia imp
related to to converting a uint4 to float4 directly:
temp1 = ((float4)(temp[i])) * one / intMax;
we have to do this for working:
((float4)(temp[i].x,temp[i].y,temp[i].z,temp[i].w))
4. quasirandom sample:
we get
:35: error: no matching overload found for arguments of type 'int, int'
int mask = pow(2, k);
^~~
:45: error: no matching overload found for arguments of type 'int, int'
output[global_id] = temp / pow(2, 32);
change you parameters to float putting (float)
pow((float)2,(float)k)
5. sobel
we get
:10: warning: unknown '#pragma OPENCL EXTENSION' - ignored
#pragma OPENCL EXTENSION cl_khr_byte_addressable_store : enable
:35: error: no matching overload found for arguments of type 'int, int'
outputImage[x + y * width] = hypot(Gx,Gy)/2;
change to:
hypot((float)Gx,(float)Gy)/2;
Resource erros:
Histogram:
genius guys at ATI are using 32k shared mem reserved to DX11 GPUs. so no Nvidia GPU can handle them (also note that with the fix below applied to 16K doesn't work also on Nvidia so seems they can't use exactly all the mem, or my error..)
For their devices it works because the local mem support is either in CPU (main mem no problem on size), 4xxx boards (emulated global mem) and 5xxx boards (uses native LDS mem which is 32kbytes )
it because a kernel in host reserves this shared mem:
BIN_SIZE*GROUP_SIZE*BIN_SIZE
which have these values:
#define BIN_SIZE 256
#define GROUP_SIZE 128
change that to:
#define BIN_SIZE 128
#define GROUP_SIZE 64
I thinks this in host and kernel code..
Mandelbrot
They are breaking Nvidia by subtle diferences in implementations..
they do are crazy thing the launch a group a 65536 threads in 1d with workgroups of one element..
I don't know if it's a problem of having more than maximum global group size, global group size in 1D, or
exceeding registers..
Fixed by changing resolution 256x256 to 128x128:
width = 256->128
It has local grups of 1 (vs warps de 32) it doesn't work by grid size..
ya que 128 i local 1
Failed erros:
Hey I have been seen all the failed errors and I don't know where are the errors..
Note they are also the most complicated ones..
I suspect they are due or due to bugs in AMD code that exhibit by different hardware (scheduling diferences) or assumptions of threads exacuting in groups (warps vs wavefront).. less
I have also learned that Nvidia OpenCL imp has no problem handling workgroups of size 4 and 1 (which are not multiple of warp size 32).. I think CUDA has also support for it..
Note also that this is very inefficient.. also on AMD GPUs with waefront of 64.. in high end (see post by Micah in AMD Stream forums..)
So finally:
Failed
======
AESEncryptDecrypt
RadixSort
SobelFilter
Final results
========
Running AESEncryptDecrypt
Failed
Running BinarySearch
Passed!
Running BinomialOption
Passed!
Running BitonicSort
Passed!
Running BlackScholes
Passed!
Running DCT
Passed!
Running DwtHaar1D
Passed!
Running EigenValue
Passed!
Running FastWalshTransform
Passed!
Running FloydWarshall
Passed!
Running HelloCL
HelloCL!
Creating a context
Getting device info
Loading and compiling CL source
Running CL program
Done
Passed!
Running Histogram
Passed!
Running Mandelbrot
Passed!
Running MatrixMultiplication
Passed!
Running MatrixTranspose
Passed!
Running MersenneTwister
Passed!
Running MonteCarloAsian
Passed!
Running NBody
Passed!
Running PrefixSum
Passed!
Running QuasiRandomSequence
Passed!
Running RadixSort
Failed
Running RecursiveGaussian
Passed!
Running Reduction
Passed!
Running ScanLargeArrays
Passed!
Running SimpleConvolution
Passed!
Running SobelFilter
Failed
0 comments:
Post a Comment