public Mandelbrot(Platform platform, int width, int height)
{
openCLPlatform = platform;
openCLDevices = openCLPlatform.QueryDevices(DeviceType.ALL);
openCLContext = openCLPlatform.CreateDefaultContext();
openCLCQ = openCLContext.CreateCommandQueue(openCLDevices[0], CommandQueueProperties.PROFILING_ENABLE);
mandelBrotProgram = openCLContext.CreateProgramWithSource(File.ReadAllText("Mandelbrot.cl"));
try
{
mandelBrotProgram.Build();
}
catch (OpenCLException)
{
string buildLog = mandelBrotProgram.GetBuildLog(openCLDevices[0]);
MessageBox.Show(buildLog, "Build error(64 bit debug sessions in vs2008 always fail like this - debug in 32 bit or use vs2010)");
Application.Exit();
}
mandelbrotKernel = mandelBrotProgram.CreateKernel("Mandelbrot");
Left = -2.0f;
Top = 2.0f;
Right = 2.0f;
Bottom = -2.0f;
BitmapWidth = width;
BitmapHeight = height;
mandelbrotMemBuffer = openCLContext.CreateBuffer((MemFlags)((long)MemFlags.WRITE_ONLY), width * height * 4, IntPtr.Zero);
}
internal static void ValidateBuild(CLProgram openclProgram, Device device, CLError error)
{
if (CLError.None != error)
{
SizeT bufferSize = SizeT.Zero;
OpenCLError.Validate(OpenCLDriver.clGetProgramBuildInfo(openclProgram, device.CLDeviceID, CLProgramBuildInfo.Log, SizeT.Zero, IntPtr.Zero, ref bufferSize));
byte[] buffer = new byte[(Int64)bufferSize];
GCHandle bufferHandle = GCHandle.Alloc(buffer, GCHandleType.Pinned);
try
{
OpenCLError.Validate(OpenCLDriver.clGetProgramBuildInfo(openclProgram, device.CLDeviceID, CLProgramBuildInfo.Log, new SizeT(buffer.LongLength), bufferHandle.AddrOfPinnedObject(), ref bufferSize));
Int32 count = Array.IndexOf <byte>(buffer, 0);
throw new OpenCLBuildError(error, System.Text.Encoding.ASCII.GetString(buffer, 0, count < 0 ? buffer.Length : count).Trim());
}
finally
{
bufferHandle.Free();
}
}
}
public TemplateMatching()
{
Device = new Device();
Context = Device.CreateContext();
CommandQueue = Context.CreateCommandQueue();
Program = Context.CreateProgram(ShaderSource);
PgKernel = Program.CreateKernel("templateMatching");
}
internal Program(CLProgram openclProgram, String[] sources, Context context, Devices devices, String buildOptions)
{
this.Context = context;
this.Devices = devices;
this.BuildOptions = buildOptions;
this.CLProgram = openclProgram;
this.Sources = new ReadOnlyIndexer <String>(sources);
}
private void SetupCLProgram()
{
if (_context != null || _kernels.Count == 0)
{
return;
}
_context = CLContext.CreateContextForDevices(CLContext.Gpus.First());
_queue = new CLCommandQueue(_context);
_program = new CLProgram(_context, _kernels.ToArray());
}
/// <summary>
/// プログラムのコンパイルまでを行います
/// </summary>
/// <param name="programSource"></param>
public static void Initialize(params string[] programSource)
{
IntPtr[] device = getDevices(getPlatforms()[0], DeviceType.Default);
Context = new CLContext(device);
CommandQueue = new CLCommandQueue(Context, device[0]);
Program = new CLProgram(Context, programSource);
Program.Build(Devices);
}
private string EmbedProgramSource(CLProgram prog, string progInit)
{
string header = $"private void {progInit}()";
string body = "string src = \"{0}\";";
return(header +
"\n{\n" +
string.Format(body, Convert.ToBase64String(Encoding.UTF8.GetBytes(prog.Source))) +
"\n}\n");
}
public static Program Create(String[] sources, Context context, Devices devices, String buildOptions)
{
CLError error = CLError.None;
CLProgram openclProgram = OpenCLDriver.clCreateProgramWithSource(context.CLContext, sources.Length, sources, GetLengths(sources), ref error);
OpenCLError.Validate(error);
OpenCLBuildError.ValidateBuild(openclProgram, devices[0], OpenCLDriver.clBuildProgram(openclProgram, devices.Count, devices.OpenCLDeviceArray, buildOptions, null, IntPtr.Zero));
return(new Program(openclProgram, sources, context, devices, buildOptions));
}
private static FLDataContainer InitializeCLKernels(string kernelPath)
{
{
CLAPI instance = CLAPI.GetInstance();
Logger.Log(LogType.Log, "Discovering Files in Path: " + kernelPath, 1);
string[] files = IOManager.DirectoryExists(kernelPath)
? IOManager.GetFiles(kernelPath, "*.cl")
: new string[0];
if (files.Length == 0)
{
Logger.Log(LogType.Error, "Error: No Files found at path: " + kernelPath, 1);
}
KernelDatabase dataBase = new KernelDatabase(DataVectorTypes.Uchar1);
List <CLProgramBuildResult> results = new List <CLProgramBuildResult>();
bool throwEx = false;
int kernelCount = 0;
int fileCount = 0;
foreach (string file in files)
{
Logger.Log(
LogType.Log,
$"[{fileCount}/{files.Length}]Loading: {file} ({kernelCount})",
2
);
try
{
CLProgram prog = dataBase.AddProgram(instance, file, false, out CLProgramBuildResult res);
kernelCount += prog.ContainedKernels.Count;
throwEx |= !res;
results.Add(res);
}
catch (Exception e)
{
Logger.Log(LogType.Error, "ERROR: " + e.Message, 2);
}
fileCount++;
}
Logger.Log(LogType.Log, "Kernels Loaded: " + kernelCount, 1);
FLInstructionSet iset = FLInstructionSet.CreateWithBuiltInTypes(dataBase);
BufferCreator creator = BufferCreator.CreateWithBuiltInTypes();
FLParser parser = new FLParser(iset, creator, new WorkItemRunnerSettings(true, 2));
return(new FLDataContainer(instance, iset, creator, parser));
}
}
public static void CreateAsync <T>(CreateCallback <T> createCallback, T userData, String[] sources, Devices devices, Context context, String buildOptions)
{
CLError error = CLError.None;
CLProgram openclProgram = OpenCLDriver.clCreateProgramWithSource(context.CLContext, sources.Length, sources, GetLengths(sources), ref error);
OpenCLError.Validate(error);
Program program = new Program(openclProgram, sources, context, devices, buildOptions);
program.AsyncHelperObject = new AsyncHelper <T>(program, createCallback, userData);
OpenCLError.Validate(OpenCLDriver.clBuildProgram(openclProgram, devices.Count, devices.OpenCLDeviceArray, buildOptions, program.AsyncCallback, IntPtr.Zero));
}
public void SignatureParsing()
{
CLAPI instance = CLAPI.GetInstance();
KernelDatabase db = new KernelDatabase(DataVectorTypes.Uchar1);
CLProgram program = db.AddProgram(instance, TEST_KERNEL, "./", true, out CLProgramBuildResult result);
CLKernel kernel = program.ContainedKernels["set_value"];
Assert.True(CheckParameter(kernel.Parameter["arr"], "arr", true, 0, DataVectorTypes.Uchar1, MemoryScope.Global));
Assert.True(CheckParameter(kernel.Parameter["value"], "value", false, 1, DataVectorTypes.Uchar1, MemoryScope.None));
db.Dispose();
instance.Dispose();
}
/// <summary>
/// Setup OpenCL
/// </summary>
internal static void Setup()
{
try
{
// ADD YOUR CODE HERE
CLError err;
uint num_entries = 0;// get platform
CLPlatformID[] platforms = new CLPlatformID[5];
err = OpenCLDriver.clGetPlatformIDs(5, platforms, ref num_entries);
if (num_entries == 0)
{
throw new Exception("No Platform Entries found!");
}
//get device ID
CLDeviceID[] devices = new CLDeviceID[1];
err = OpenCLDriver.clGetDeviceIDs(platforms[0], CLDeviceType.GPU, 1, devices, ref num_entries);
// create context
ctx = OpenCLDriver.clCreateContext(null, 1, devices, null, IntPtr.Zero, ref err);
if (err != CLError.Success)
{
throw new Exception(err.ToString());
}
// create command queue
comQ = OpenCLDriver.clCreateCommandQueue(ctx, devices[0], 0, ref err);
if (err != CLError.Success)
{
throw new Exception(err.ToString());
}
// Compile Program
string[] progString = new string[1];
progString[0] = File.ReadAllText("..\\..\\prog.cl");
program = OpenCLDriver.clCreateProgramWithSource(ctx, 1, progString, null, ref err);
err = OpenCLDriver.clBuildProgram(program, 1, devices, "", null, IntPtr.Zero);
//create kernel
kernel_mult = OpenCLDriver.clCreateKernel(program, "multiply", ref err);
}
catch (Exception exc)
{
MessageBox.Show(exc.ToString());
}
}
public void Dispose()
{
if (!(PgKernel is null))
{
PgKernel.Dispose();
PgKernel = null;
}
if (!(Program is null))
{
Program.Dispose();
Program = null;
}
if (!(CommandQueue is null))
{
CommandQueue.Dispose();
CommandQueue = null;
}
if (!(Context is null))
{
Context.Dispose();
Context = null;
}
}
/// <summary>
/// Setup OpenCL
/// </summary>
internal static void Setup()
{
try
{
// ADD YOUR CODE HERE
CLError err;
uint num_entries = 0;// get platform
CLPlatformID[] platforms = new CLPlatformID[5];
err = OpenCLDriver.clGetPlatformIDs(5, platforms, ref num_entries);
if (num_entries == 0) throw new Exception("No Platform Entries found!");
//get device ID
CLDeviceID[] devices = new CLDeviceID[1];
err = OpenCLDriver.clGetDeviceIDs(platforms[0], CLDeviceType.GPU, 1, devices, ref num_entries);
// create context
ctx = OpenCLDriver.clCreateContext(null, 1, devices, null, IntPtr.Zero, ref err);
if (err != CLError.Success) throw new Exception(err.ToString());
// create command queue
comQ = OpenCLDriver.clCreateCommandQueue(ctx, devices[0], 0, ref err);
if (err != CLError.Success) throw new Exception(err.ToString());
// Compile Program
string[] progString = new string[1];
progString[0] = File.ReadAllText("..\\..\\prog.cl");
program = OpenCLDriver.clCreateProgramWithSource(ctx, 1, progString, null, ref err);
err = OpenCLDriver.clBuildProgram(program, 1, devices, "", null, IntPtr.Zero);
//create kernel
kernel_mult = OpenCLDriver.clCreateKernel(program, "multiply", ref err);
}
catch (Exception exc)
{
MessageBox.Show(exc.ToString());
}
}
public static void ConvertToGrayscale(string inputPath)
{
CLResultCode error = CLResultCode.Success;
//Get a platform
CLPlatform[] platforms = new CLPlatform[1];
CL.GetPlatformIds(1, platforms, out _);
//Get a device
CLDevice[] devices = new CLDevice[1];
CL.GetDeviceIds(platforms[0], DeviceType.Gpu, 1, devices, out _);
//Create a context
CLContext context = CL.CreateContext(IntPtr.Zero, devices, IntPtr.Zero, IntPtr.Zero, out error);
if (error != CLResultCode.Success)
{
throw new Exception("Error on creating a context");
}
//Create the program from source
CLProgram program =
CL.CreateProgramWithSource(context, File.ReadAllText("Kernels/grayscale.cl"), out error);
error = CL.BuildProgram(program, 1, devices, null, IntPtr.Zero, IntPtr.Zero);
if (error != CLResultCode.Success)
{
throw new Exception($"Error on building program: {error}");
}
//Get the kernel which we will use
CLKernel kernel = CL.CreateKernel(program, "grayscale", out error);
ImageFormat inputImageFormat = new ImageFormat
{
ChannelOrder = ChannelOrder.Bgra,
ChannelType = ChannelType.UnsignedInteger8
};
Bitmap inputBitmap = new Bitmap(inputPath);
BitmapData inputData = inputBitmap.LockBits(new Rectangle(0, 0, inputBitmap.Width, inputBitmap.Height),
ImageLockMode.ReadOnly, PixelFormat.Format32bppArgb);
ImageDescription imageDescription = new ImageDescription()
{
ImageType = MemoryObjectType.Image2D,
Width = (UIntPtr)inputBitmap.Width,
Height = (UIntPtr)inputBitmap.Height,
Depth = (UIntPtr)1
};
CLImage inputImage = CL.CreateImage(context, MemoryFlags.ReadOnly | MemoryFlags.CopyHostPtr,
ref inputImageFormat, ref imageDescription, inputData.Scan0, out error);
Bitmap outputBitmap = new Bitmap(inputBitmap.Width, inputBitmap.Height, PixelFormat.Format32bppArgb);
BitmapData outputData = outputBitmap.LockBits(new Rectangle(0, 0, inputBitmap.Width, inputBitmap.Height),
ImageLockMode.WriteOnly, PixelFormat.Format32bppArgb);
CLImage outputImage = CL.CreateImage(context, MemoryFlags.WriteOnly | MemoryFlags.UseHostPtr,
ref inputImageFormat, ref imageDescription, outputData.Scan0, out error);
CL.SetKernelArg(kernel, 0, rFact);
CL.SetKernelArg(kernel, 1, gFact);
CL.SetKernelArg(kernel, 2, bFact);
CL.SetKernelArg(kernel, 3, inputImage);
CL.SetKernelArg(kernel, 4, outputImage);
CLCommandQueue queue = CL.CreateCommandQueueWithProperties(context, devices[0], IntPtr.Zero, out error);
CL.EnqueueNDRangeKernel(queue, kernel, 2, new UIntPtr[] { UIntPtr.Zero, UIntPtr.Zero, }, new[]
{
(UIntPtr)inputBitmap.Width,
(UIntPtr)inputBitmap.Height,
}, null, 0, null, out _);
CL.EnqueueReadImage(queue, outputImage, 1, new UIntPtr[] { UIntPtr.Zero, UIntPtr.Zero, },
new UIntPtr[]
{
(UIntPtr)inputBitmap.Width,
(UIntPtr)inputBitmap.Height,
(UIntPtr)1,
}, UIntPtr.Zero, UIntPtr.Zero, outputData.Scan0, 0, null, out _);
outputBitmap.UnlockBits(outputData);
outputBitmap.Save("grayscale.png", System.Drawing.Imaging.ImageFormat.Png);
//Dispose of all things
inputBitmap.Dispose();
outputBitmap.Dispose();
CL.ReleaseMemoryObject(inputImage);
CL.ReleaseMemoryObject(outputImage);
CL.ReleaseCommandQueue(queue);
CL.ReleaseKernel(kernel);
CL.ReleaseProgram(program);
CL.ReleaseContext(context);
}
private void AsyncCallback(CLProgram program, IntPtr user_data)
{
AsyncHelperObject.Callback();
AsyncHelperObject = null;
}
private void Initialize()
{
CLAPI instance = CLAPI.GetInstance();
string path = FLScriptEditor.Settings.KernelPath;
StartupSequence.loaderForm.SetStatus("Discovering Files in Path: " + path);
string[] files = IOManager.DirectoryExists(path) ? IOManager.GetFiles(path, "*.cl") : new string[0];
if (files.Length == 0)
{
DialogResult res = StyledMessageBox.Show(
"Error",
"No Files found at path: " + path,
MessageBoxButtons.AbortRetryIgnore,
SystemIcons.Error
);
if (res == DialogResult.Retry)
{
Initialize();
return;
}
if (res == DialogResult.Abort)
{
StartupSequence.loaderForm.DialogResult = DialogResult.Abort;
StartupSequence.loaderForm.Close();
return;
}
if (res == DialogResult.Ignore)
{
StartupSequence.loaderForm.DialogResult = DialogResult.OK;
}
}
KernelDatabase dataBase = new KernelDatabase(DataVectorTypes.Uchar1);
List <CLProgramBuildResult> results = new List <CLProgramBuildResult>();
bool throwEx = false;
int kernelCount = 0;
int fileCount = 0;
if (FLScriptEditor.Settings.ExperimentalKernelLoading)
{
try
{
string source = TextProcessorAPI.PreprocessSource(files, new Dictionary <string, bool>());
CLProgram prog = dataBase.AddProgram(instance, source, "./", false, out CLProgramBuildResult res);
throwEx |= !res;
if (res)
{
kernelCount += prog.ContainedKernels.Count;
}
results.Add(res);
StartupSequence.loaderForm.SetStatus($"File Loaded(Kernels Loaded): ({kernelCount})");
}
catch (Exception e)
{
throw new SoftException(e);
}
}
else
{
foreach (string file in files)
{
StartupSequence.loaderForm.SetStatus(
$"[{fileCount}/{files.Length}]Loading: {file} ({kernelCount})"
);
try
{
CLProgram prog = dataBase.AddProgram(instance, file, false, out CLProgramBuildResult res);
kernelCount += prog.ContainedKernels.Count;
throwEx |= !res;
results.Add(res);
}
catch (Exception e)
{
StartupSequence.loaderForm.Log("ERROR: " + e.Message, Color.Red);
throw e; //Let the Exception Viewer Catch that
}
fileCount++;
}
}
StartupSequence.loaderForm.SetStatus("Loading Finished");
StartupSequence.loaderForm.Log("Loading Finished", Color.White);
StartupSequence.loaderForm.Log("Kernels Loaded: " + kernelCount, Color.White);
if (throwEx)
{
DialogResult res =
StyledMessageBox.Show(
"OpenCL Build Errors",
"There are errors in one or more OpenCL kernels. Do you want to open the OpenCL Build Excepion Viewer?",
MessageBoxButtons.YesNoCancel,
SystemIcons.Warning
);
if (res == DialogResult.Cancel)
{
StartupSequence.loaderForm.DialogResult = DialogResult.Abort;
StartupSequence.loaderForm.Close();
}
else if (res == DialogResult.Yes)
{
BuildExceptionViewer bvr = new BuildExceptionViewer(new CLBuildException(results));
if (bvr.ShowDialog() == DialogResult.Retry)
{
dataBase.Dispose();
Initialize();
}
}
}
FLInstructionSet iset = FLInstructionSet.CreateWithBuiltInTypes(dataBase);
BufferCreator creator = BufferCreator.CreateWithBuiltInTypes();
FLParser parser = new FLParser(iset, creator, new WorkItemRunnerSettings(true, 2));
StartupSequence.FlContainer = new FLDataContainer(instance, iset, creator, parser);
}
public static void Add()
{
CLResultCode error = CLResultCode.Success;
CLPlatform[] platforms = new CLPlatform[1];
CL.GetPlatformIds(1, platforms, out _);
CLDevice[] devices = new CLDevice[1];
CL.GetDeviceIds(platforms[0], DeviceType.All, 1, devices, out _);
CLContext context = CL.CreateContext(IntPtr.Zero, devices, IntPtr.Zero, IntPtr.Zero, out _);
CLProgram program =
CL.CreateProgramWithSource(context, File.ReadAllText("Kernels/add_arrays.cl"), out error);
error = CL.BuildProgram(program, 1, devices, null, IntPtr.Zero, IntPtr.Zero);
if (error != CLResultCode.Success)
{
throw new Exception(error.ToString());
}
CLKernel kernel = CL.CreateKernel(program, "add", out error);
Span <float> inputA = new float[]
{
1, 24, 5, 43, 41, 56
};
Span <float> inputB = new float[]
{
72, -323, -1, 43, -41, -26
};
Span <float> output = stackalloc float[inputA.Length];
CLBuffer bufferA = CL.CreateBuffer(context, MemoryFlags.ReadOnly | MemoryFlags.CopyHostPtr, inputA, out _);
CLBuffer bufferB = CL.CreateBuffer(context, MemoryFlags.ReadOnly | MemoryFlags.CopyHostPtr, inputB, out _);
CLBuffer outputBuffer = CL.CreateBuffer(context, MemoryFlags.WriteOnly, (UIntPtr)(output.Length * sizeof(float)), IntPtr.Zero, out _);
//For outputs I wouldn't use that also enqueueing a ReadBuffer is needed regardless
//CLBuffer outputBuffer = CL.CreateBuffer(context, MemoryFlags.WriteOnly | MemoryFlags.UseHostPtr, output, out _);
CL.SetKernelArg(kernel, 0, bufferA);
CL.SetKernelArg(kernel, 1, bufferB);
CL.SetKernelArg(kernel, 2, outputBuffer);
CLCommandQueue queue = CL.CreateCommandQueueWithProperties(context, devices[0], IntPtr.Zero, out error);
CL.EnqueueNDRangeKernel(queue, kernel, 1, null,
new[] { (UIntPtr)inputA.Length }, null, 0, null, out _);
CL.EnqueueReadBuffer(queue, outputBuffer, true, UIntPtr.Zero, output, null, out _);
foreach (float f in output)
{
Console.WriteLine(f);
}
CL.ReleaseMemoryObject(bufferA);
CL.ReleaseMemoryObject(bufferB);
CL.ReleaseMemoryObject(outputBuffer);
CL.ReleaseCommandQueue(queue);
CL.ReleaseKernel(kernel);
CL.ReleaseProgram(program);
CL.ReleaseContext(context);
}
static void Main(string[] args)
{
ToGrayscale.ConvertToGrayscale("image.jpg");
//Get the ids of available opencl platforms
CL.GetPlatformIds(0, null, out uint platformCount);
CLPlatform[] platformIds = new CLPlatform[platformCount];
CL.GetPlatformIds(platformCount, platformIds, out _);
Console.WriteLine(platformIds.Length);
foreach (CLPlatform platform in platformIds)
{
Console.WriteLine(platform.Handle);
CL.GetPlatformInfo(platform, PlatformInfo.Name, out byte[] val);
}
//Get the device ids for each platform
foreach (IntPtr platformId in platformIds)
{
CL.GetDeviceIds(new CLPlatform(platformId), DeviceType.All, out CLDevice[] deviceIds);
CLContext context = CL.CreateContext(IntPtr.Zero, (uint)deviceIds.Length, deviceIds, IntPtr.Zero,
IntPtr.Zero, out CLResultCode result);
if (result != CLResultCode.Success)
{
throw new Exception("The context couldn't be created.");
}
string code = @"
__kernel void add(__global float* A, __global float* B,__global float* result, const float mul)
{
int i = get_global_id(0);
result[i] = (A[i] + B[i])*mul;
}";
CLProgram program = CL.CreateProgramWithSource(context, code, out result);
CL.BuildProgram(program, (uint)deviceIds.Length, deviceIds, null, IntPtr.Zero, IntPtr.Zero);
CLKernel kernel = CL.CreateKernel(program, "add", out result);
int arraySize = 20;
float[] A = new float[arraySize];
float[] B = new float[arraySize];
for (int i = 0; i < arraySize; i++)
{
A[i] = 1;
B[i] = i;
}
CLBuffer bufferA = CL.CreateBuffer(context, MemoryFlags.ReadOnly | MemoryFlags.CopyHostPtr, A,
out result);
CLBuffer bufferB = CL.CreateBuffer(context, MemoryFlags.ReadOnly | MemoryFlags.CopyHostPtr, B,
out result);
float[] pattern = new float[] { 1, 3, 5, 7 };
CLBuffer resultBuffer = new CLBuffer(CL.CreateBuffer(context, MemoryFlags.WriteOnly,
new UIntPtr((uint)(arraySize * sizeof(float))), IntPtr.Zero, out result));
try
{
CL.SetKernelArg(kernel, 0, bufferA);
CL.SetKernelArg(kernel, 1, bufferB);
CL.SetKernelArg(kernel, 2, resultBuffer);
CL.SetKernelArg(kernel, 3, -1f);
CLCommandQueue commandQueue = new CLCommandQueue(
CL.CreateCommandQueueWithProperties(context, deviceIds[0], IntPtr.Zero, out result));
CL.EnqueueFillBuffer(commandQueue, bufferB, pattern, UIntPtr.Zero, (UIntPtr)(arraySize * sizeof(float)), null,
out _);
//CL.EnqueueNDRangeKernel(commandQueue, kernel, 1, null, new UIntPtr[] {new UIntPtr((uint)A.Length)},
// null, 0, null, out CLEvent eventHandle);
CL.EnqueueNDRangeKernel(commandQueue, kernel, 1, null, new UIntPtr[] { new UIntPtr((uint)A.Length) },
null, 0, null, out CLEvent eventHandle);
CL.Finish(commandQueue);
CL.SetEventCallback(eventHandle, (int)CommandExecutionStatus.Complete, (waitEvent, data) =>
{
float[] resultValues = new float[arraySize];
CL.EnqueueReadBuffer(commandQueue, resultBuffer, true, UIntPtr.Zero, resultValues, null, out _);
StringBuilder line = new StringBuilder();
foreach (float res in resultValues)
{
line.Append(res);
line.Append(", ");
}
Console.WriteLine(line.ToString());
});
//get rid of the buffers because we no longer need them
CL.ReleaseMemoryObject(bufferA);
CL.ReleaseMemoryObject(bufferB);
CL.ReleaseMemoryObject(resultBuffer);
//Release the program kernels and queues
CL.ReleaseProgram(program);
CL.ReleaseKernel(kernel);
CL.ReleaseCommandQueue(commandQueue);
CL.ReleaseContext(context);
CL.ReleaseEvent(eventHandle);
}
catch (Exception e)
{
Console.WriteLine(e.ToString());
throw;
}
}
}
public Kernel CreateKernel(string source, string kernelName)
{
var program = new CLProgram(source, this);
return(program.CreateKernel(kernelName));
}
private (string, string, CLProgram[]) GenerateTargets(SerializableFLInstruction[] targets)
{
string newName = "opt";
string newSig = "";
List <string> lines = new List <string>();
Dictionary <string, string> funcs = new Dictionary <string, string>();
List <CLProgram> progs = new List <CLProgram>();
int count = 0;
foreach (SerializableFLInstruction serializableFlInstruction in targets)
{
CLKernel instr = InstructionSet.Database.GetClKernel(serializableFlInstruction.InstructionKey);
CLProgram prog = InstructionSet.Database.GetProgram(instr);
progs.Add(prog);
newName += "_" + instr.Name;
string funcSig = "";
List <(string orig, string newKey)> reps = new List <(string orig, string newKey)>();
foreach (KeyValuePair <string, KernelParameter> kernelParameter in instr.Parameter.Skip(5))
{
funcSig += ", ";
switch (kernelParameter.Value.MemScope)
{
case MemoryScope.None:
break;
case MemoryScope.Global:
funcSig += "__global ";
break;
case MemoryScope.Constant:
funcSig += "__constant ";
break;
default:
throw new ArgumentOutOfRangeException();
}
funcSig += KernelParameter.GetDataString(kernelParameter.Value.DataType);
funcSig += kernelParameter.Value.IsArray ? "* " : " ";
funcSig += kernelParameter.Key + "_mrge_" + count;
reps.Add((kernelParameter.Key, kernelParameter.Key + "_mrge_" + count));
count++;
}
newSig += funcSig;
string[] block = GetBlockContent(prog.Source, instr.Name);
foreach ((string orig, string newKey)valueTuple in reps)
{
for (int i = 0; i < block.Length; i++)
{
int current = block[i].IndexOf(valueTuple.orig, StringComparison.Ordinal);
while (current != -1)
{
if (CheckBack(block[i], current + valueTuple.orig.Length - 1) &&
CheckFront(block[i], current))
{
block[i] = block[i].Remove(current, valueTuple.orig.Length)
.Insert(current, valueTuple.newKey);
}
current = block[i].IndexOf(
valueTuple.orig,
current + valueTuple.orig.Length,
StringComparison.Ordinal
);
}
}
}
string genFuncArgs = reps.Count == 0 ? "" : ", " + reps.Select(x => x.newKey).Unpack(", ");
if (!funcs.ContainsKey(instr.Name))
{
funcs.Add(
instr.Name,
GetFunc(
"gen_" + instr.Name,
"__global uchar* image, int3 dimensions, int channelCount, float maxValue, __global uchar* channelEnableState" +
funcSig,
block
)
);
}
string line =
$"\ngen_{instr.Name}(image, dimensions, channelCount, maxValue, channelEnableState{genFuncArgs});\n";
lines.Add(line);
}
string sig =
$"__kernel void {newName}(__global uchar* image, int3 dimensions, int channelCount, float maxValue, __global uchar* channelEnableState{newSig})";
string newk = $"{funcs.Values.Distinct().Unpack("\n\n")}\n{sig}\n" + "{" + lines.Unpack("\n") + "\n}";
return(newName, newk, progs.ToArray());
}
