def _may_have_svm(dev):
    has_svm = (dev.platform._get_cl_version() >= (2, 0)
            and cl.get_cl_header_version() >= (2, 0))

    if dev.platform.name == "Portable Computing Language":
        has_svm = (
                get_pocl_version(dev.platform) >= (1, 0)
                and cl.get_cl_header_version() >= (2, 0))

    return has_svm

def get_info():
    global selected_device, selected_platform, context, KERNELS_DEFS
    info = {
            "version.cl_header"     : pyopencl.get_cl_header_version(),
            "opengl"                : pyopencl.have_gl(),
            #"kernels"               : KERNELS_DEFS.keys()
            }
    updict(info, "pyopencl", get_pyopencl_info())
    if selected_platform:
        updict(info, "platform", {
            "name"          : selected_platform.name,
            "vendor"        : selected_platform.vendor,
            "devices"       : len(selected_platform.get_devices()),
            })
    if selected_device:
        if hasattr(selected_device, "opencl_c_version"):
            info["device.opencl_c_version"] = getattr(selected_device, "opencl_c_version")
        updict(info, "device", {
            "type"                      : device_type(selected_device),
            "name"                      : selected_device.name.strip(),
            "version"                   : selected_device.version,
            "max_work_group_size"       : selected_device.max_work_group_size,
            "max_work_item_dimensions"  : selected_device.max_work_item_dimensions,
            "max_work_item_sizes"       : selected_device.max_work_item_sizes,
            "max-size"                  : selected_device_max_size})
    return info

def get_info():
    global selected_device, selected_platform, context, KERNELS_DEFS
    info = {"version"               : pyopencl.version.VERSION,
            "version.text"          : pyopencl.VERSION_TEXT,
            "version.status"        : pyopencl.VERSION_STATUS,
            "version.cl_header"     : pyopencl.get_cl_header_version(),
            "opengl"                : pyopencl.have_gl(),
            #"kernels"               : KERNELS_DEFS.keys()
            }
    if selected_platform:
        info.update({
            "platform.name"         : selected_platform.name,
            "platform.vendor"       : selected_platform.vendor,
            "platform.devices"      : len(selected_platform.get_devices()),
            })
    if selected_device:
        if hasattr(selected_device, "opencl_c_version"):
            info["device.opencl_c_version"] = getattr(selected_device, "opencl_c_version")
        info.update({
            "device.type"           : device_type(selected_device),
            "device.name"           : selected_device.name.strip(),
            "device.version"        : selected_device.version,
            "device.max_work_group_size"        : selected_device.max_work_group_size,
            "device.max_work_item_dimensions"   : selected_device.max_work_item_dimensions,
            "device.max_work_item_sizes"        : selected_device.max_work_item_sizes})
    return info

def test_custom_type_zeros(ctx_factory):
    context = ctx_factory()
    queue = cl.CommandQueue(context)

    if not (
            queue._get_cl_version() >= (1, 2)
            and cl.get_cl_header_version() >= (1, 2)):
        pytest.skip("CL1.2 not available")

    dtype = np.dtype([
        ("cur_min", np.int32),
        ("cur_max", np.int32),
        ("pad", np.int32),
        ])

    from pyopencl.tools import get_or_register_dtype, match_dtype_to_c_struct

    name = "mmc_type"
    dtype, c_decl = match_dtype_to_c_struct(queue.device, name, dtype)
    dtype = get_or_register_dtype(name, dtype)

    n = 1000
    z_dev = cl.array.zeros(queue, n, dtype=dtype)

    z = z_dev.get()

    assert np.array_equal(np.zeros(n, dtype), z)

def test_sub_buffers(ctx_factory):
    ctx = ctx_factory()
    if (ctx._get_cl_version() < (1, 1) or
            cl.get_cl_header_version() < (1, 1)):
        from pytest import skip
        skip("sub-buffers are only available in OpenCL 1.1")

    alignment = ctx.devices[0].mem_base_addr_align

    queue = cl.CommandQueue(ctx)

    n = 30000
    a = (np.random.rand(n) * 100).astype(np.uint8)

    mf = cl.mem_flags
    a_buf = cl.Buffer(ctx, mf.READ_WRITE | mf.COPY_HOST_PTR, hostbuf=a)

    start = (5000 // alignment) * alignment
    stop = start + 20 * alignment

    a_sub_ref = a[start:stop]

    a_sub = np.empty_like(a_sub_ref)
    cl.enqueue_copy(queue, a_sub, a_buf[start:stop])

    assert np.array_equal(a_sub, a_sub_ref)

def test_platform_get_devices(ctx_factory):
    ctx = ctx_factory()
    platform = ctx.devices[0].platform

    if platform.name == "Apple":
        pytest.xfail("Apple doesn't understand all the values we pass " "for dev_type")

    dev_types = [
        cl.device_type.ACCELERATOR,
        cl.device_type.ALL,
        cl.device_type.CPU,
        cl.device_type.DEFAULT,
        cl.device_type.GPU,
    ]
    if (
        platform._get_cl_version() >= (1, 2)
        and cl.get_cl_header_version() >= (1, 2)
        and not platform.name.lower().startswith("nvidia")
    ):
        dev_types.append(cl.device_type.CUSTOM)

    for dev_type in dev_types:
        print(dev_type)
        devs = platform.get_devices(dev_type)
        if dev_type in (cl.device_type.DEFAULT, cl.device_type.ALL, getattr(cl.device_type, "CUSTOM", None)):
            continue
        for dev in devs:
            assert dev.type & dev_type == dev_type

def test_unload_compiler(platform):
    if (platform._get_cl_version() < (1, 2) or
            cl.get_cl_header_version() < (1, 2)):
        from pytest import skip
        skip("clUnloadPlatformCompiler is only available in OpenCL 1.2")
    _skip_if_pocl(platform, 'pocl does not support unloading compiler')
    if platform.vendor == "Intel(R) Corporation":
        from pytest import skip
        skip("Intel proprietary driver does not support unloading compiler")
    cl.unload_platform_compiler(platform)

def test_fine_grain_svm(ctx_factory):
    ctx = ctx_factory()
    # queue = cl.CommandQueue(ctx)

    if (ctx._get_cl_version() < (2, 0) or
            cl.get_cl_header_version() < (2, 0)):
        from pytest import skip
        skip("SVM only available in OpenCL 2.0 and higher")

    svm_ary = cl.fsvm_empty(ctx, (100, 100), np.float32, alignment=64)
    assert isinstance(svm_ary.base, cl.SVMAllocation)

def test_coarse_grain_svm(ctx_factory):
    import sys
    is_pypy = '__pypy__' in sys.builtin_module_names

    ctx = ctx_factory()
    queue = cl.CommandQueue(ctx)

    if (ctx._get_cl_version() < (2, 0) or
            cl.get_cl_header_version() < (2, 0)):
        from pytest import skip
        skip("SVM only available in OpenCL 2.0 and higher")
    dev = ctx.devices[0]
    if ("AMD" in dev.platform.name
            and dev.type & cl.device_type.CPU):
        pytest.xfail("AMD CPU doesn't do coarse-grain SVM")

    n = 3000
    svm_ary = cl.SVM(cl.csvm_empty(ctx, (n,), np.float32, alignment=64))
    if not is_pypy:
        # https://bitbucket.org/pypy/numpy/issues/52
        assert isinstance(svm_ary.mem.base, cl.SVMAllocation)

    if dev.platform.name != "Portable Computing Language":
        # pocl 0.13 has a bug misinterpreting the size parameter
        cl.enqueue_svm_memfill(queue, svm_ary, np.zeros((), svm_ary.mem.dtype))

    with svm_ary.map_rw(queue) as ary:
        ary.fill(17)
        orig_ary = ary.copy()

    prg = cl.Program(ctx, """
        __kernel void twice(__global float *a_g)
        {
          a_g[get_global_id(0)] *= 2;
        }
        """).build()

    prg.twice(queue, svm_ary.mem.shape, None, svm_ary)

    with svm_ary.map_ro(queue) as ary:
        print(ary)
        assert np.array_equal(orig_ary*2, ary)

    new_ary = np.empty_like(orig_ary)
    new_ary.fill(-1)

    if ctx.devices[0].platform.name != "Portable Computing Language":
        # "Blocking memcpy is unimplemented (clEnqueueSVMMemcpy.c:61)"
        # in pocl 0.13.

        cl.enqueue_copy(queue, new_ary, svm_ary)
        assert np.array_equal(orig_ary*2, new_ary)

def test_platform_get_devices(platform):
    dev_types = [cl.device_type.ACCELERATOR, cl.device_type.ALL,
                 cl.device_type.CPU, cl.device_type.DEFAULT, cl.device_type.GPU]
    if (platform._get_cl_version() >= (1, 2) and
            cl.get_cl_header_version() >= (1, 2)):
        dev_types.append(cl.device_type.CUSTOM)
    for dev_type in dev_types:
        devs = platform.get_devices(dev_type)
        if dev_type in (cl.device_type.DEFAULT,
                        cl.device_type.ALL,
                        getattr(cl.device_type, 'CUSTOM', None)):
            continue
        for dev in devs:
            assert dev.type == dev_type

def test_user_event(ctx_factory):
    ctx = ctx_factory()
    if ctx._get_cl_version() < (1, 1) and cl.get_cl_header_version() < (1, 1):
        from pytest import skip

        skip("UserEvent is only available in OpenCL 1.1")

    if ctx.devices[0].platform.name == "Portable Computing Language":
        # https://github.com/pocl/pocl/issues/201
        pytest.xfail("POCL's user events don't work right")

    status = {}

    def event_waiter1(e, key):
        e.wait()
        status[key] = True

    def event_waiter2(e, key):
        cl.wait_for_events([e])
        status[key] = True

    from threading import Thread
    from time import sleep

    evt = cl.UserEvent(ctx)
    Thread(target=event_waiter1, args=(evt, 1)).start()
    sleep(0.05)
    if status.get(1, False):
        raise RuntimeError("UserEvent triggered before set_status")
    evt.set_status(cl.command_execution_status.COMPLETE)
    sleep(0.05)
    if not status.get(1, False):
        raise RuntimeError("UserEvent.wait timeout")
    assert evt.command_execution_status == cl.command_execution_status.COMPLETE

    evt = cl.UserEvent(ctx)
    Thread(target=event_waiter2, args=(evt, 2)).start()
    sleep(0.05)
    if status.get(2, False):
        raise RuntimeError("UserEvent triggered before set_status")
    evt.set_status(cl.command_execution_status.COMPLETE)
    sleep(0.05)
    if not status.get(2, False):
        raise RuntimeError("cl.wait_for_events timeout on UserEvent")
    assert evt.command_execution_status == cl.command_execution_status.COMPLETE

def test_user_event(ctx_factory):
    ctx = ctx_factory()
    if (ctx._get_cl_version() < (1, 1) and
            cl.get_cl_header_version() < (1, 1)):
        from pytest import skip
        skip("UserEvent is only available in OpenCL 1.1")

    status = {}

    def event_waiter1(e, key):
        e.wait()
        status[key] = True

    def event_waiter2(e, key):
        cl.wait_for_events([e])
        status[key] = True

    from threading import Thread
    from time import sleep
    evt = cl.UserEvent(ctx)
    Thread(target=event_waiter1, args=(evt, 1)).start()
    sleep(.05)
    if status.get(1, False):
        raise RuntimeError('UserEvent triggered before set_status')
    evt.set_status(cl.command_execution_status.COMPLETE)
    sleep(.05)
    if not status.get(1, False):
        raise RuntimeError('UserEvent.wait timeout')
    assert evt.command_execution_status == cl.command_execution_status.COMPLETE

    evt = cl.UserEvent(ctx)
    Thread(target=event_waiter2, args=(evt, 2)).start()
    sleep(.05)
    if status.get(2, False):
        raise RuntimeError('UserEvent triggered before set_status')
    evt.set_status(cl.command_execution_status.COMPLETE)
    sleep(.05)
    if not status.get(2, False):
        raise RuntimeError('cl.wait_for_events timeout on UserEvent')
    assert evt.command_execution_status == cl.command_execution_status.COMPLETE

def test_fine_grain_svm(ctx_factory):
    import sys
    is_pypy = '__pypy__' in sys.builtin_module_names

    ctx = ctx_factory()
    queue = cl.CommandQueue(ctx)

    from pytest import skip
    if (ctx._get_cl_version() < (2, 0) or
            cl.get_cl_header_version() < (2, 0)):
        skip("SVM only available in OpenCL 2.0 and higher")

    if not (ctx.devices[0].svm_capabilities
            & cl.device_svm_capabilities.FINE_GRAIN_BUFFER):
        skip("device does not support fine-grain SVM")

    n = 3000
    ary = cl.fsvm_empty(ctx, n, np.float32, alignment=64)

    if not is_pypy:
        # https://bitbucket.org/pypy/numpy/issues/52
        assert isinstance(ary.base, cl.SVMAllocation)

    ary.fill(17)
    orig_ary = ary.copy()

    prg = cl.Program(ctx, """
        __kernel void twice(__global float *a_g)
        {
          a_g[get_global_id(0)] *= 2;
        }
        """).build()

    prg.twice(queue, ary.shape, None, cl.SVM(ary))
    queue.finish()

    print(ary)
    assert np.array_equal(orig_ary*2, ary)

def test_spirv(ctx_factory):
    ctx = ctx_factory()
    queue = cl.CommandQueue(ctx)

    if (ctx._get_cl_version() < (2, 1) or
            cl.get_cl_header_version() < (2, 1)):
        from pytest import skip
        skip("SPIR-V program creation only available in OpenCL 2.1 and higher")

    n = 50000

    a_dev = cl.clrandom.rand(queue, n, np.float32)
    b_dev = cl.clrandom.rand(queue, n, np.float32)
    dest_dev = cl_array.empty_like(a_dev)

    with open("add-vectors-%d.spv" % queue.device.address_bits, "rb") as spv_file:
        spv = spv_file.read()

    prg = cl.Program(ctx, spv)

    prg.sum(queue, a_dev.shape, None, a_dev.data, b_dev.data, dest_dev.data)

    assert la.norm((dest_dev - (a_dev+b_dev)).get()) < 1e-7

def test_get_info(ctx_factory):
    ctx = ctx_factory()
    device, = ctx.devices
    platform = device.platform

    failure_count = [0]

    pocl_quirks = [
        (cl.Buffer, cl.mem_info.OFFSET),
        (cl.Program, cl.program_info.BINARIES),
        (cl.Program, cl.program_info.BINARY_SIZES),
    ]
    if ctx._get_cl_version() >= (1, 2) and cl.get_cl_header_version() >= (1, 2):
        pocl_quirks.extend([
            (cl.Program, cl.program_info.KERNEL_NAMES),
            (cl.Program, cl.program_info.NUM_KERNELS),
        ])
    CRASH_QUIRKS = [  # noqa
            (("NVIDIA Corporation", "NVIDIA CUDA",
                "OpenCL 1.0 CUDA 3.0.1"),
                [
                    (cl.Event, cl.event_info.COMMAND_QUEUE),
                    ]),
            (("The pocl project", "Portable Computing Language",
                "OpenCL 1.2 pocl 0.8-pre"),
                    pocl_quirks),
            (("The pocl project", "Portable Computing Language",
                "OpenCL 1.2 pocl 0.8"),
                pocl_quirks),
            (("The pocl project", "Portable Computing Language",
                "OpenCL 1.2 pocl 0.9-pre"),
                pocl_quirks),
            (("The pocl project", "Portable Computing Language",
                "OpenCL 1.2 pocl 0.9"),
                pocl_quirks),
            (("The pocl project", "Portable Computing Language",
                "OpenCL 1.2 pocl 0.10-pre"),
                pocl_quirks),
            (("The pocl project", "Portable Computing Language",
                "OpenCL 1.2 pocl 0.10"),
                pocl_quirks),
            (("Apple", "Apple",
                "OpenCL 1.2"),
                [
                    (cl.Program, cl.program_info.SOURCE),
                    ]),
            ]
    QUIRKS = []  # noqa

    def find_quirk(quirk_list, cl_obj, info):
        for (vendor, name, version), quirks in quirk_list:
            if (
                    vendor == platform.vendor
                    and name == platform.name
                    and platform.version.startswith(version)):
                for quirk_cls, quirk_info in quirks:
                    if (isinstance(cl_obj, quirk_cls)
                            and quirk_info == info):
                        return True

        return False

    def do_test(cl_obj, info_cls, func=None, try_attr_form=True):
        if func is None:
            def func(info):
                cl_obj.get_info(info)

        for info_name in dir(info_cls):
            if not info_name.startswith("_") and info_name != "to_string":
                print(info_cls, info_name)
                info = getattr(info_cls, info_name)

                if find_quirk(CRASH_QUIRKS, cl_obj, info):
                    print("not executing get_info", type(cl_obj), info_name)
                    print("(known crash quirk for %s)" % platform.name)
                    continue

                try:
                    func(info)
                except:
                    msg = "failed get_info", type(cl_obj), info_name

                    if find_quirk(QUIRKS, cl_obj, info):
                        msg += ("(known quirk for %s)" % platform.name)
                    else:
                        failure_count[0] += 1

                if try_attr_form:
                    try:
                        getattr(cl_obj, info_name.lower())
                    except:
                        print("failed attr-based get_info", type(cl_obj), info_name)

                        if find_quirk(QUIRKS, cl_obj, info):
                            print("(known quirk for %s)" % platform.name)
                        else:
                            failure_count[0] += 1

    do_test(platform, cl.platform_info)
    do_test(device, cl.device_info)
#.........这里部分代码省略.........

import pprint
from idxread import idxs
from npsolve import runner
from math import ceil, log2

pp = pprint.PrettyPrinter(depth=5)
mf = cl.mem_flags

def dec2str(num):
    k = []
    s = str(num)
    for a in s:
        k.append(ascii_lowercase[int(a)])
    return 'qq'+''.join(k)

print( cl.get_cl_header_version() )

ctx = cl.create_some_context()
queue = cl.CommandQueue(ctx)

set = 'random'
traindata = idxs("train-images-idx3-ubyte.idx", "train-labels-idx1-ubyte.idx")
testdata  = idxs("t10k-images-idx3-ubyte.idx", "t10k-labels-idx1-ubyte.idx")
result = 1.0
ninpt =  traindata.count                 #Samples count ( 60000 for set )
nvarsd = traindata.rows*traindata.cols   #Count of equations members ( 28*28 for set)
topology = [nvarsd, 5, 4, 3, 1]
nvarsg = genn.countcns(topology)     #Count of equations members
print("Total connections is", nvarsg)
nsamp = 64#ctx.get_info(cl.context_info.DEVICES)[0].max_work_group_size #Genome samples count (current sort limitation to local_size)
print("Population count is", nsamp)

def log_version_info():
    log.info("PyOpenCL loaded, header version: %s, GL support: %s",
             ".".join([str(x) for x in pyopencl.get_cl_header_version()]), pyopencl.have_gl())

'''
Prints relevant information regarding the capabilities of the current OpenCL runtime and devices
Note that pyopencl has a script that prints all properties in its examples folder
'''

import pyopencl as cl

print('PyOpenCL version: ' + cl.VERSION_TEXT)
print('OpenCL header version: ' + '.'.join(map(str, cl.get_cl_header_version())) + '\n')

# Get installed platforms (SDKs)
print('- Installed platforms (SDKs) and available devices:')
platforms = cl.get_platforms()

for plat in platforms:
    indent = ''

    # Get and print platform info
    print(indent + '{} ({})'.format(plat.name, plat.vendor))
    indent = '\t'
    print(indent + 'Version: ' + plat.version)
    print(indent + 'Profile: ' + plat.profile)
    print(indent + 'Extensions: ' + str(plat.extensions.strip().split(' ')))

    # Get and print device info
    devices = plat.get_devices(cl.device_type.ALL)

    print(indent + 'Available devices: ')
    if not devices:
        print(indent + '\tNone')