Any example of clCreateProgramWithBinary use

[davidcastells]

I'm working with OpenCL on FPGAs, so I would like to load OpenCL programs from binary files.

All the examples I've seen (maybe I'm missing something) are creating programs from source. But I see that the code is somehow prepared to create programs from binary files.

Could you provide an example of how to do so?

BR

[inducer]

Following

https://documen.tician.de/pyopencl/runtime_program.html#pyopencl.Program

you can just do

prg = cl.Program(ctx, (device,), (binary,))

where binary is the bytes object containing the binary.

[inducer]

Run with gdb --args python ...?

[davidcastells]

It looks like it happens when the Altera (now Intel) code tries to free something. This does not happen in C/C++ code.
I still have to try with Xilinx bitstreams..

Thread 1 "python" received signal SIGSEGV, Segmentation fault.
__GI___libc_free (mem=0x7c046a007c047c03) at malloc.c:3113
3113    malloc.c: No such file or directory.
(gdb) bt
#0  __GI___libc_free (mem=0x7c046a007c047c03) at malloc.c:3113
#1  0x00007fffccc1fbd9 in l_load_binary_pkg () from /opt/intelFPGA/17.1/hld/host/linux64/lib/libalteracl.so
#2  0x00007fffccc2167a in clCreateProgramWithBinaryIntelFPGA () from /opt/intelFPGA/17.1/hld/host/linux64/lib/libalteracl.so
#3  0x00007ffff40c5bf3 in clCreateProgramWithBinary (context=0x158ae20, num_devices=1, device_list=0x15e2310, lengths=0x15e2750, binaries=0x15e2730,
    binary_status=<optimized out>, errcode_ret=0x7fffffffc3b8) at ocl_icd_loader_gen.c:2032
#4  0x00007ffff418f40f in pyopencl::create_program_with_binary(pyopencl::context&, pybind11::sequence, pybind11::sequence) ()
   from /home/dcr/.local/lib/python3.7/site-packages/pyopencl/_cl.cpython-37m-x86_64-linux-gnu.so
#5  0x00007ffff415d018 in void pybind11::cpp_function::initialize<void pybind11::detail::initimpl::factory<pyopencl_expose_part_2(pybind11::module_&)::{lambda(pyopencl::context&, pybind11::sequence, pybind11::sequence)#6}, pybind11::detail::void_type (*)(), pyopencl::program* (pyopencl::context&, pybind11::sequence, pybind11::sequence), pybind11::detail::void_type>::execute<pybind11::class_<pybind11::detail::void_type (*)()>, pybind11::arg, pybind11::class_<pybind11::detail::void_type (*)()>, pybind11::class_<pybind11::detail::void_type (*)()> >(pybind11::class_<pybind11::detail::void_type (*)()>&, pybind11::arg const&, pybind11::class_<pybind11::detail::void_type (*)()> const&, pybind11::class_<pybind11::detail::void_type (*)()> const&) &&::{lambda(pybind11::detail::value_and_holder&, pyopencl::context&, pybind11::sequence, pybind11::sequence)#1}, void, pybind11::detail::value_and_holder, pyopencl::context&, pybind11::sequence, pybind11::sequence, pybind11::name, pybind11::is_method, pybind11::sibling, pybind11::detail::is_new_style_constructor, pybind11::class_<pybind11::detail::void_type (*)()>, pybind11::class_<pybind11::detail::void_type (*)()>, pybind11::class_<pybind11::detail::void_type (*)()> >(pybind11::arg&&, void (*)(pybind11::detail::value_and_holder, pyopencl::context&, pybind11::sequence, pybind11::sequence), pybind11::name const&, pybind11::is_method const&, pybind11::sibling const&, pybind11::detail::is_new_style_constructor const&, pybind11::class_<pybind11::detail::void_type (*)()> const&, pybind11::class_<pybind11::detail::void_type (*)()> const&, pybind11::class_<pybind11::detail::void_type (*)()> const&)::{lambda(pybind11::detail::function_call&)#3}::_FUN ()
   from /home/dcr/.local/lib/python3.7/site-packages/pyopencl/_cl.cpython-37m-x86_64-linux-gnu.so
#6  0x00007ffff411bd0f in pybind11::cpp_function::dispatcher(_object*, _object*, _object*) ()
   from /home/dcr/.local/lib/python3.7/site-packages/pyopencl/_cl.cpython-37m-x86_64-linux-gnu.so
#7  0x00000000005bc266 in _PyMethodDef_RawFastCallDict () at Objects/call.c:515
#8  0x00000000004c310a in _PyCFunction_FastCallDict (kwargs=0x0, nargs=4, args=0x7fffffffc7f0, func=0x7ffff4086e10) at Objects/call.c:582
#9  _PyObject_FastCallDict (kwargs=0x0, nargs=4, args=0x7fffffffc7f0, callable=0x7ffff4086e10) at Objects/call.c:101
#10 _PyObject_Call_Prepend () at Objects/call.c:904
#11 0x00000000004c3786 in PyObject_Call () at Objects/call.c:245
#12 0x000000000043bd49 in slot_tp_init (self=0x7ffff3fde150, args=0x7ffff4086480, kwds=0x0) at Objects/typeobject.c:6613
#13 0x0000000000501d19 in type_call () at Objects/typeobject.c:949
#14 0x00007ffff411a0cb in pybind11_meta_call () from /home/dcr/.local/lib/python3.7/site-packages/pyopencl/_cl.cpython-37m-x86_64-linux-gnu.so
#15 0x00000000004c17fa in _PyObject_FastCallKeywords () at Objects/call.c:199
#16 0x0000000000538e46 in call_function (kwnames=0x0, oparg=<optimized out>, pp_stack=<synthetic pointer>) at Python/ceval.c:4619
#17 _PyEval_EvalFrameDefault () at Python/ceval.c:3093
#18 0x0000000000532bb5 in PyEval_EvalFrameEx (throwflag=0, f=0x140a328) at Python/ceval.c:547
#19 _PyEval_EvalCodeWithName () at Python/ceval.c:3930
#20 0x00000000004c3464 in _PyFunction_FastCallDict (kwargs=0x0, nargs=4, args=0x7fffffffcc10, func=0x7ffff4069158) at Objects/call.c:376
#21 _PyObject_FastCallDict (kwargs=0x0, nargs=4, args=0x7fffffffcc10, callable=0x7ffff4069158) at Objects/call.c:98
#22 _PyObject_Call_Prepend () at Objects/call.c:904
#23 0x0000000000504da7 in slot_tp_init () at Objects/typeobject.c:6610
#24 0x0000000000501d19 in type_call () at Objects/typeobject.c:949
#25 0x00000000004c17fa in _PyObject_FastCallKeywords () at Objects/call.c:199
#26 0x0000000000538e46 in call_function (kwnames=0x0, oparg=<optimized out>, pp_stack=<synthetic pointer>) at Python/ceval.c:4619
#27 _PyEval_EvalFrameDefault () at Python/ceval.c:3093
#28 0x0000000000532bb5 in PyEval_EvalFrameEx (throwflag=0, f=0x15f0c78) at Python/ceval.c:547
#29 _PyEval_EvalCodeWithName () at Python/ceval.c:3930
#30 0x00000000004c2ac3 in _PyFunction_FastCallKeywords () at Objects/call.c:433
#31 0x00000000005383bc in call_function (kwnames=0x0, oparg=<optimized out>, pp_stack=<synthetic pointer>) at Python/ceval.c:4616
#32 _PyEval_EvalFrameDefault () at Python/ceval.c:3093
#33 0x0000000000532bb5 in PyEval_EvalFrameEx (throwflag=0, f=0x992ab8) at Python/ceval.c:547
#34 _PyEval_EvalCodeWithName () at Python/ceval.c:3930
---Type <return> to continue, or q <return> to quit---
#35 0x0000000000532a63 in PyEval_EvalCodeEx (closure=0x0, kwdefs=0x0, defcount=0, defs=0x0, kwcount=0, kws=0x0, argcount=0, args=0x0, locals=locals@entry=0x7ffff7ecc1f8,
    globals=globals@entry=0x7ffff7ecc1f8, _co=_co@entry=0x7ffff7f05660) at Python/ceval.c:3959
#36 PyEval_EvalCode (co=co@entry=0x7ffff7f05660, globals=globals@entry=0x7ffff7ecc1f8, locals=locals@entry=0x7ffff7ecc1f8) at Python/ceval.c:524
#37 0x00000000005de0f2 in run_mod () at Python/pythonrun.c:1035
#38 0x00000000005de52a in PyRun_FileExFlags () at Python/pythonrun.c:988
#39 0x00000000005de317 in PyRun_SimpleFileExFlags () at Python/pythonrun.c:429
#40 0x00000000005b70f1 in pymain_run_file (p_cf=0x7fffffffd2b0, filename=0x94f840 L"perf_close3d_oclfpga_closingperpixel.py", fp=0x997360) at Modules/main.c:427
#41 pymain_run_filename (cf=0x7fffffffd2b0, pymain=0x7fffffffd3c0) at Modules/main.c:1627
#42 pymain_run_python (pymain=0x7fffffffd3c0) at Modules/main.c:2876
#43 pymain_main () at Modules/main.c:3037
#44 0x00000000005b6dba in _Py_UnixMain () at Modules/main.c:3072
#45 0x00007ffff703fbf7 in __libc_start_main (main=0x4bd810 <main>, argc=2, argv=0x7fffffffd518, init=<optimized out>, fini=<optimized out>, rtld_fini=<optimized out>,
    stack_end=0x7fffffffd508) at ../csu/libc-start.c:310
#46 0x00000000005b6ca9 in _start ()

[inducer]

This looks like an issue with the Altera ICD, not sure that PyOpenCL could do to avoid this issue.

[davidcastells]

I doesn't look like.
I've discovered that the error is different when using lists or tuples

This code
program = cl.Program(context, [device], [binary])
....raises the free error and crash the app

and this code
program = cl.Program(context, (device,), (binary,))
...raises and INVALID_BINARY error exiting in a controlled way

Still a similar code in (C/C++) works perfectly

I get this behaviour in pyopencl versions <= 2020.1

pyopencl versions >= 2020.2 don't work with my FPGA because it seems that pyopencl is not supporting OpenCL versions < 2.1

ImportError: /opt/intelFPGA/17.1/hld/host/linux64/lib/libOpenCL.so.1: version `OPENCL_2.1' not found (required by /home/dcr/.local/lib/python3.7/site-packages/pyopencl/_cl.cpython-37m-x86_64-linux-gnu.so)

Is this so? Or otherwise, can I fix it somehow?

[davidcastells]

OK. I solved the issue 🎉.
Last pyopencl version is working.

And the solution is 🥁🥁🥁.... removing a "import matplotlib.pyplot" that I had

Not joking, apparently it makes pyopencl on my fpga crash. Don't ask my why, it's really wierd

[davidcastells]

Let me provide more info.
I've been trying to trigger the same error from a C++ test program without success.

The program prints SUCCESS when done right, and raises a segmentation fault when I introduce invalid parameters, but never in free from libc (as we get when called from PyOpenCL).

I attach the small code snippet with comments about the tests I did.

Any idea ?

 cl_device_id devices[]={did};
 size_t fileLen;
 unsigned char* pFile = loadBinaryFile("close_nd.aocx", &fileLen);

 const unsigned char* binaries[]={pFile};

 size_t binary_lengths[]={fileLen};
 cl_int binary_status[1];   
 // context = (cl_context) 3;  // THIS CAUSES A SEGMENTATION FAULT before entering the altera library
 // cl_int* pErr = (cl_int*) 5; // THIS CAUSES A SEGMENTATION FAULT in the altera library, but not in free
 // cl_int* binary_status = (cl_int*) 4; // THIS CAUSES A SEGMENTATION FAULT in the altera library, but not in free
 // size_t* binary_lengths = (size_t*) 4; // THIS CAUSES A SEGMENTATION FAULT in the altera library, but not in free
 // const unsigned char* binaries[] = {(unsigned char*) 64};  // THIS CAUSES A SEGMENTATION FAULT in the altera library when trying to copy (not in free)
 //  const unsigned char** binaries = (const unsigned char**) 4;  // THIS CAUSES A SEGMENTATION FAULT in the altera library, but not in free

 cl_program program = clCreateProgramWithBinary(context, 1, devices, binary_lengths, binaries, binary_status, &err);
 assert(err == CL_SUCCESS);
 printf("SUCCESS\n");

[yves-surrel]

It happens that I just came into that problem.

On my Mac MBP, the following code:

#!/usr/bin/env python3
# -*- coding: utf-8 -*-
"""
Created on Sun Mar 13 09:58:27 2022

@author: yves
"""

import os
import pyopencl as cl

# Kernel function
src = """
__kernel void test(
    __global int *a)
{
    a[get_global_id(0)] = 0;
}
"""

platforms = cl.get_platforms()

# Save binaries
for i_platform, platform in enumerate(platforms):
    devices = platform.get_devices()
    for i_device, device in enumerate(devices):
        ctx = cl.Context([device])
        prg = cl.Program(ctx, src).build()
        binary, = prg.get_info(cl.program_info.BINARIES)
        print(prg.get_info(cl.program_info.BINARY_SIZES)[0])
        filename = f'test_{i_platform}_{i_device}.bin'
        if os.path.exists(filename):
            os.remove(filename)
        with open(filename, 'xb') as bin:
            bin.write(binary)
            

# Read binaries and create programs
for i_platform, platform in enumerate(platforms):
    devices = platform.get_devices()
    for i_device, device in enumerate(devices):
        ctx = cl.Context([device])
        with open(f'test_{i_platform}_{i_device}.bin', 'rb') as bin:
            binary = bin.read()
            print(len(binary))
            prg = cl.Program(ctx, [device], [binary])
            print(prg.get_info(cl.program_info.BINARY_SIZES)[0])
            try:
                print(prg.all_kernels())
            except cl.LogicError as e:
                print(e)

returns

223
1796
13659
223
0
clCreateKernelsInProgram failed: INVALID_PROGRAM_EXECUTABLE
1796
0
clCreateKernelsInProgram failed: INVALID_PROGRAM_EXECUTABLE
13659
0
clCreateKernelsInProgram failed: INVALID_PROGRAM_EXECUTABLE

I should be missing something...

[davidcastells]

In your example you compile and save the binary file from pyopencl, and then program the binary file from pyopencl as well. As it fails, we don't know wether the error is from the compilation phase or the programming from binary phase.

Can you separe building from programming ? and test that you are able to program the binary from a C/C++ executable?
By the way, can you try if using tuples instead of lists make any difference?

prg=cl.Program(ctx,(device,),(binary,))

[yves-surrel]

It's the same with prg = cl.Program(ctx, (device,), (binary,)) (contrary would have been surprising...)

I do not understand what you mean by 'separe building from programming'. What kind of change in the source do you expect?

My example should be runnable on your computer (Minimal Working Example). How does it behave?

[yves-surrel]

...and I am not a C++ programmer

[yves-surrel]

An even smaller working example (without writing files):

import os
import pyopencl as cl

# Kernel function
src = """
__kernel void test(
    __global int *a)
{
    a[get_global_id(0)] = 0;
}
"""

platforms = cl.get_platforms()

for platform in platforms:
    devices = platform.get_devices()
    for i_device, device in enumerate(devices):
        ctx = cl.Context([device])
        prg = cl.Program(ctx, src).build()
        binary, = prg.get_info(cl.program_info.BINARIES)
        # Create Program from binary
        prg1 = cl.Program(ctx, (device,), (binary,))
        try:
            print(prg1.all_kernels())
        except cl.LogicError as e:
            print(e)

[yves-surrel]

On my BootCamp Windows, it simply crashes python :-(

[yves-surrel]

I have found somewhere in StackOverflow: " if you want to save a compiled kernel, save it after a call to clCompileProgram method. On #Mojave you cannot build a program from binary if that binary was previously built by clBuildProgram". Seems true.
If I replace buildby compile, it works:

import os
import pyopencl as cl

# Kernel function
src = """
__kernel void test(
    __global int *a)
{
    a[get_global_id(0)] = 0;
}
"""

platforms = cl.get_platforms()

for platform in platforms:
    devices = platform.get_devices()
    for i_device, device in enumerate(devices):
        ctx = cl.Context([device])
        prg = cl.Program(ctx, src).compile()
        sizes, = prg.get_info(cl.program_info.BINARY_SIZES)
        binary, = prg.get_info(cl.program_info.BINARIES)
        # Create Program from binary
        prg1 = cl.Program(ctx, (device,), (binary,))
        prg1.build()
        for knl in prg1.all_kernels():
            print(knl.function_name)

prints

test
test
test
/Volumes/FAT/Developpement/Python/core/untitled6.py:27: UserWarning: Pre-build attribute access defeats compiler caching.
  prg = cl.Program(ctx, src).compile()

I do not know how to suppress the warning (and the relevance of the warning either).

[inducer]

prg.get_info(cl.program_info.BINARY_SIZES) can be prg.binary_sizes. (etc.)

I do not know how to suppress the warning (and the relevance of the warning either).

The warnings mean that the PyOpenCL compiler cache won't be used. If you're not waiting longer than you're willing for things to compile, then it's safe to ignore.