An even more polished example, with full source:
Code:
test.cl:
Code:
test2.cl:
Code:
This test shows off the Compute library running using both OpenCL computing and forcing to the CPU fallback. The first test simply copies a string and inserts different text based on which compute function is used. The second test simply prints the local ID (stays at zero as the local workers are mucked about with internally):
Code:
To show off the fallback in action, here is what happens if the OpenCL code fails (the kernel fails to compile in this case):
Code:
Oh, and did I mention that all of the output is styled/colored?
Random sidenote: Current class count of the project is 91 classes with 16 subprojects.
Code:
// Main test case
#include <context.h>
#include <task.h>
#include <taskworker.h>
#include <iostream>
#include <string.h>
#include <ffi.h>
#include <colors.h>
using namespace std;
using namespace Compute;
// Fallback for test 1
void testCPU(TaskWorker* wrk, char* in, char* out)
{
const char name[] = "CPU";
unsigned int i = wrk->get_global_id(0);
if(i >= 28 && i <= 30)
{
out[i] = name[i - 28];
}
else
{
out[i] = in[i];
}
}
// Fallback for test 2
void testCPU2(TaskWorker* wrk, unsigned int* out)
{
unsigned int g = wrk->get_global_id(0);
unsigned int l = wrk->get_local_id(0);
out[g] = l;
}
int main(int argc, char** argv)
{
string input;
char output[256];
auto test1 = [&input, &output](bool forceCPU)
{
// Clear out output
memset(output, 0, sizeof(output));
// Make Task, Context is auto-created as Singleton
Task test("Test task", "test.cl", true, "main", testCPU);
// Fallback?
if(forceCPU) test.forceFallback();
// Input arguments
test.setArgument(input.size() + 1, (void*)input.data(), CL_MEM_COPY_HOST_PTR | CL_MEM_READ_ONLY, &ffi_type_pointer);
// Output arguments
test.setReturn(256, output, CL_MEM_WRITE_ONLY, &ffi_type_pointer);
// Dimensions
test.setWorkItems(1, input.size() + 1);
// Run it
test.start();
test.join();
test.end();
// Check results
cout << output << endl;
};
// Normal test
input = string("Hello world from AHelper0's ___ 550 Ti from OpenCL");
test1(false);
// Fallback test
input = string("Hello world from AHelper0's ___ using libffi in fallback mode");
test1(true);
unsigned int locals[10];
auto test2 = [&locals](bool forceCPU)
{
Task test2("Test2 task", "test2.cl", true, "main", testCPU2);
test2.setReturn(sizeof(locals), locals, CL_MEM_WRITE_ONLY, &ffi_type_pointer);
test2.setWorkItems(1, 10);
if(forceCPU)
test2.forceFallback(); // Forcing can happen anywhere, even mid-OpenCL setup
test2.start();
test2.join();
test2.end();
for(int i = 0; i < 10; i++)
{
cout << locals[i] << " ";
}
cout << endl;
};
// Normal test
test2(false);
// Fallback test
test2(true); // 0's because local mode not supported. Default is to say 0 on error in OpenCL, so...
}
test.cl:
Code:
__kernel void main(__global char *in, __global char *out)
{
const char name[] = "GTX";
unsigned int i = get_global_id(0);
if(i >= 28 && i <= 30)
{
out[i] = name[i-28];
}
else
{
out[i] = in[i];
}
}
test2.cl:
Code:
__kernel void main(__global unsigned int *out)
{
unsigned int g = get_global_id(0);
unsigned int l = get_local_id(0);
out[g] = l;
}
This test shows off the Compute library running using both OpenCL computing and forcing to the CPU fallback. The first test simply copies a string and inserts different text based on which compute function is used. The second test simply prints the local ID (stays at zero as the local workers are mucked about with internally):
Code:
![HAL@HAL build]$ ./test5
Compute::Compute(): Loaded libOpenCL.so
Compute::Compute(): Detected 1 device(s):
GeForce GTX 550 Ti 1024
Compute library initialized
Hello world from AHelper0's GTX 550 Ti from OpenCL
Hello world from AHelper0's CPU using libffi in fallback mode
0 1 2 3 4 5 6 7 8 9
0 0 0 0 0 0 0 0 0 0
Compute library unloaded
To show off the fallback in action, here is what happens if the OpenCL code fails (the kernel fails to compile in this case):
Code:
![HAL@HAL build]$ ./test5
Compute::Compute(): Loaded libOpenCL.so
Compute::Compute(): Detected 1 device(s):
GeForce GTX 550 Ti 1024
Compute library initialized
Compute::Task::Task(): Failed to compile CL code: :9:14: error: use of undeclared identifier 'namE'; did you mean 'name'?
out[i] = namE[i-28];
^~~~
name
:3:14: note: 'name' declared here
const char name[] = "GTX";
^
Hello world from AHelper0's CPU 550 Ti from OpenCL
Compute::Task::Task(): Failed to compile CL code: :9:14: error: use of undeclared identifier 'namE'; did you mean 'name'?
out[i] = namE[i-28];
^~~~
name
:3:14: note: 'name' declared here
const char name[] = "GTX";
^
Hello world from AHelper0's CPU using libffi in fallback mode
0 1 2 3 4 5 6 7 8 9
0 0 0 0 0 0 0 0 0 0
Compute library unloaded
Oh, and did I mention that all of the output is styled/colored?
Random sidenote: Current class count of the project is 91 classes with 16 subprojects.