diff --git a/doc/make_constants.py b/doc/make_constants.py
index d5b444a5e6668ebf5ae863b866c75924e2f212a6..e45a064acc707587e3175b5c229e09cbadd73b31 100644
--- a/doc/make_constants.py
+++ b/doc/make_constants.py
@@ -4,11 +4,17 @@ def doc_class(cls):
     print ".. class :: %s" % cls.__name__
     print
     for i in sorted(dir(cls)):
-        if not i.startswith("_"):
+        if not i.startswith("_")  and not i == "to_string":
             print "    .. attribute :: %s" % i
+    print "    .. method :: to_string(value)"
+    print
+    print "        Returns a :class:`str` representing *value*."
+    print
+    print "        .. versionadded:: 0.91"
     print
 
 
 print ".. This is an automatically generated file. DO NOT EDIT"
+print
 for cls in cl.CONSTANT_CLASSES:
     doc_class(cls)
diff --git a/doc/source/misc.rst b/doc/source/misc.rst
index 61abb7fb5bcd1cce980b0041eea4a28a376ea799..a07d964eb69733f3b60c4be90e83f141b8058d93 100644
--- a/doc/source/misc.rst
+++ b/doc/source/misc.rst
@@ -7,12 +7,39 @@ Installation information is maintained collaboratively on the
 Acknowledgments
 ===============
 
-* James Snyder provided a patch to make PyOpenCL work on OS X 10.6.
+* James Snyder provided patches to make PyOpenCL work on OS X 10.6.
 * Roger Pau Monné supplied the example :file:`examples/benchmark-all.py`.
 
 User-visible Changes
 ====================
 
+Version 0.91
+------------
+
+* Add a test suite.
+* Fix numerous `get_info` bugs. (reports by David Garcia and the test suite)
+* Add :meth:`pyopencl.ImageFormat.__repr__`.
+* Add :meth:`pyopencl.addressing_mode.to_string` and colleagues.
+* The `pitch` arguments to 
+  :func:`pyopencl.create_image_2d`,
+  :func:`pyopencl.create_image_3d`,
+  :func:`pyopencl.enqueue_read_image`, and
+  :func:`pyopencl.enqueue_write_image`
+  are now defaulted to zero. The argument order of `enqueue_{read,write}_image`
+  has changed for this reason.
+* :meth:`pyopencl.MemoryObject.get_image_info` now actually exists.
+* Add :meth:`pyopencl.MemoryObject.image`.
+
+Version 0.90.4
+--------------
+
+* Add build fixes for Windows and OS X.
+
+Version 0.90.3
+--------------
+
+* Fix a GNU-ism in the C++ code of the wrapper.
+
 Version 0.90.2
 --------------
 
diff --git a/doc/source/reference.rst b/doc/source/reference.rst
index 8ea5f616ebd5087feced671ee2c6f3c2a550631a..fd0b8049133b3d3c7c3d19818725d6f0b0de9ff1 100644
--- a/doc/source/reference.rst
+++ b/doc/source/reference.rst
@@ -188,15 +188,24 @@ Memory
 
         See :class:`mem_info` for values of *param*.
 
+    .. method:: get_image_info(param)
+
+        See :class:`image_info` for values of *param*.
+
     .. attribute:: info
 
         Lower case versions of the :class:`mem_info` constants
         may be used as attributes on instances of this class
         to directly query info attributes.
 
-    .. method:: get_image_info(param)
+    .. attribute:: image.info
 
-        See :class:`image_info` for values of *param*.
+        Lower case versions of the :class:`image_info` constants
+        may be used as attributes on the attribute `image` of this 
+        class to directly query image info.
+
+        For example, you may use *img.image.depth* instead of
+        *img.get_image_info(pyopencl.image_info.DEPTH)*.
 
     .. method:: release()
 
@@ -237,6 +246,12 @@ Image Formats
 
         See :class:`channel_type` for possible values.
 
+    .. method:: __repr__
+
+        Returns a :class:`str` representation of the image format.
+
+        .. versionadded:: 0.91
+
 .. function:: get_supported_image_formats(context, flags, image_type)
 
     See :class:`mem_flags` for possible values of *flags*
@@ -245,21 +260,21 @@ Image Formats
 Images
 ^^^^^^
 
-.. function:: create_image_2d(context, flags, format, width, height, pitch, host_buffer=None)
+.. function:: create_image_2d(context, flags, format, width, height, pitch=0, host_buffer=None)
 
     See :class:`mem_flags` for possible values of *flags*.
     Returns a new image-type :class:`MemoryObject`.
 
-.. function:: create_image_3d(context, flags, format, width, height, depth, row_pitch, slice_pitch, host_buffer=None)
+.. function:: create_image_3d(context, flags, format, width, height, depth, row_pitch=0, slice_pitch=0, host_buffer=None)
 
     See :class:`mem_flags` for possible values of *flags*.
     Returns a new image-type :class:`MemoryObject`.
 
-.. function:: enqueue_read_image(queue, mem, origin, region, row_pitch, slice_pitch, host_buffer, wait_for=None, is_blocking=False)
+.. function:: enqueue_read_image(queue, mem, origin, region, host_buffer, row_pitch=0, slice_pitch=0, wait_for=None, is_blocking=False)
 
     |enqueue-waitfor|
 
-.. function:: enqueue_write_image(queue, mem, origin, region, row_pitch, slice_pitch, host_buffer, wait_for=None, is_blocking=False)
+.. function:: enqueue_write_image(queue, mem, origin, region, host_buffer, row_pitch=0, slice_pitch=0, wait_for=None, is_blocking=False)
 
     |enqueue-waitfor|
 
@@ -331,7 +346,7 @@ Programs and Kernels
         may be used as attributes on instances of this class
         to directly query info attributes.
 
-    .. method:: get_build_info(param, device)
+    .. method:: get_build_info(device, param)
 
         See :class:`program_build_info` for values of *param*.
 
@@ -353,6 +368,10 @@ Programs and Kernels
 
     |comparable|
 
+    .. method:: all_kernels()
+
+        Returns a list of all :class:`Kernel` objects in the :class:`Program`.
+
 .. function:: unload_compiler()
 .. function:: create_program_with_source(context, src)
 .. function:: create_program_with_binary(context, devices, binaries)
diff --git a/examples/benchmark-all.py b/examples/benchmark-all.py
index 493dd48eb3f3b0f8895612537d8220beed35d001..93685006dc8817307c12025d5813f9c5485f1e6d 100644
--- a/examples/benchmark-all.py
+++ b/examples/benchmark-all.py
@@ -23,12 +23,6 @@ print "Execution time of test without OpenCL: ", time2 - time1, "s"
 
 for platform in cl.get_platforms():
     for device in platform.get_devices():
-        dev_type = "unknown"
-
-        for dev_type_candidate in dir(cl.device_type):
-            if getattr(cl.device_type, dev_type_candidate) == device.type:
-                dev_type = dev_type_candidate
-
         print "==============================================================="
         print "Platform name:", platform.name
         print "Platform profile:", platform.profile
@@ -36,7 +30,7 @@ for platform in cl.get_platforms():
         print "Platform version:", platform.version
         print "---------------------------------------------------------------"
         print "Device name:", device.name
-        print "Device type: ", dev_type
+        print "Device type:", cl.device_type.to_string(device.type)
         print "Device memory: ", device.global_mem_size//1024//1024, 'MB'
         print "Device max clock speed:", device.max_clock_frequency, 'MHz'
         print "Device compute units:", device.max_compute_units
diff --git a/pyopencl/__init__.py b/pyopencl/__init__.py
index a8a992051ab6aa51c8e763eef326b7762dcd7df6..2f1e03c6c91cd0cac2df56001b455108180447c5 100644
--- a/pyopencl/__init__.py
+++ b/pyopencl/__init__.py
@@ -20,6 +20,42 @@ def _add_functionality():
             _cl.Kernel: _cl.kernel_info,
             }
 
+    def to_string(cls, value):
+        for name in dir(cls):
+            if (not name.startswith("_") and getattr(cls, name) == value):
+                return name
+
+        raise ValueError("a name for value %d was not found in %s"
+                % (value, cls.__name__))
+
+    addressing_mode.to_string = classmethod(to_string)
+    channel_order.to_string = classmethod(to_string)
+    channel_type.to_string = classmethod(to_string)
+    command_execution_status.to_string = classmethod(to_string)
+    command_queue_info.to_string = classmethod(to_string)
+    command_queue_properties.to_string = classmethod(to_string)
+    context_info.to_string = classmethod(to_string)
+    context_properties.to_string = classmethod(to_string)
+    device_exec_capabilities.to_string = classmethod(to_string)
+    device_fp_config.to_string = classmethod(to_string)
+    device_info.to_string = classmethod(to_string)
+    device_local_mem_type.to_string = classmethod(to_string)
+    device_mem_cache_type.to_string = classmethod(to_string)
+    device_type.to_string = classmethod(to_string)
+    event_info.to_string = classmethod(to_string)
+    filter_mode.to_string = classmethod(to_string)
+    image_info.to_string = classmethod(to_string)
+    kernel_info.to_string = classmethod(to_string)
+    kernel_work_group_info.to_string = classmethod(to_string)
+    map_flags.to_string = classmethod(to_string)
+    mem_info.to_string = classmethod(to_string)
+    mem_object_type.to_string = classmethod(to_string)
+    platform_info.to_string = classmethod(to_string)
+    profiling_info.to_string = classmethod(to_string)
+    program_build_info.to_string = classmethod(to_string)
+    program_info.to_string = classmethod(to_string)
+    sampler_info.to_string = classmethod(to_string)
+
     class ProfilingInfoGetter:
         def __init__(self, event):
             self.event = event
@@ -37,6 +73,23 @@ def _add_functionality():
 
     _cl.Event.profile = property(ProfilingInfoGetter)
 
+    class ImageInfoGetter:
+        def __init__(self, mem):
+            self.mem = mem
+
+        def __getattr__(self, name):
+            info_cls = _cl.image_info
+
+            try:
+                inf_attr = getattr(info_cls, name.upper())
+            except AttributeError:
+                raise AttributeError("%s has no attribute '%s'"
+                        % (type(self), name))
+            else:
+                return self.mem.get_image_info(inf_attr)
+
+    _cl.MemoryObject.image = property(ImageInfoGetter)
+
     def make_getattr(info_cls):
         def result(self, name):
             try:
@@ -57,7 +110,11 @@ def _add_functionality():
             pi_attr = getattr(program_info, attr.upper())
         except AttributeError:
             try:
-                return Kernel(self, attr)
+                knl = Kernel(self, attr)
+                # Nvidia does not raise errors even for invalid names,
+                # but this will give an error if the kernel is invalid.
+                knl.num_args
+                return knl
             except LogicError:
                 raise AttributeError("'%s' was not found as a program info attribute or as a kernel name"
                         % attr)
@@ -84,6 +141,13 @@ def _add_functionality():
 
     Kernel.__call__ = kernel_call
 
+    def image_format_repr(self):
+        return "ImageFormat(%s, %s)" % (
+                channel_order.to_string(self.channel_order),
+                channel_type.to_string(self.channel_data_type))
+
+    ImageFormat.__repr__ = image_format_repr
+
     def event_wait(self):
         wait_for_events([self])
         return self
diff --git a/pyopencl/version.py b/pyopencl/version.py
index 2a05b570d53fe7b2cf6771e12d4c4edf23ff9cb3..5abc8d8af0686b44410279fda6687f4fb017ed26 100644
--- a/pyopencl/version.py
+++ b/pyopencl/version.py
@@ -1,5 +1,5 @@
-VERSION = (0, 90, 3)
-VERSION_STATUS = ""
+VERSION = (0, 91)
+VERSION_STATUS = "alpha"
 VERSION_TEXT = ".".join(str(x) for x in VERSION) + VERSION_STATUS
 
 
diff --git a/setup.py b/setup.py
index 9098a3fed908b6cc47142a7db6ca2a5d8d15a84a..c2e72f7ef6246b70917620557189f818eb203ab1 100644
--- a/setup.py
+++ b/setup.py
@@ -132,6 +132,7 @@ def main():
 
             install_requires=[
                 "pytools>=7",
+                "py>=1.0.2"
                 ],
 
             ext_package="pyopencl",
diff --git a/src/wrapper/wrap_cl.cpp b/src/wrapper/wrap_cl.cpp
index a0e05b7ea7aeb4c13f570f9dc75a2eb3c9bc6511..5a7d8c993ebbb7e048498a4984d290450b081b64 100644
--- a/src/wrapper/wrap_cl.cpp
+++ b/src/wrapper/wrap_cl.cpp
@@ -468,6 +468,7 @@ BOOST_PYTHON_MODULE(_cl)
     typedef memory_object cls;
     py::class_<cls, boost::noncopyable>("MemoryObject", py::no_init)
       .DEF_SIMPLE_METHOD(get_info)
+      .DEF_SIMPLE_METHOD(get_image_info)
       .DEF_SIMPLE_METHOD(release)
       .def(py::self == py::self)
       .def(py::self != py::self)
@@ -503,24 +504,30 @@ BOOST_PYTHON_MODULE(_cl)
 
   DEF_SIMPLE_FUNCTION(get_supported_image_formats);
   py::def("create_image_2d", create_image_2D,
-      (py::args("context", "flags", "format", "width", "height", "pitch"), 
-       py::arg("host_buffer")=py::object()),
+      (py::args("context", "flags", "format", "width", "height"), 
+       py::arg("pitch")=0,
+       py::arg("host_buffer")=py::object()
+       ),
       py::return_value_policy<py::manage_new_object>());
   py::def("create_image_3d", create_image_3D,
-      (py::args("context", "flags", "format", "width", "height", "depth", 
-                "row_pitch", "slice_pitch"), 
-       py::arg("host_buffer")=py::object()),
+      (py::args("context", "flags", "format", "width", "height", "depth"),
+       py::arg("row_pitch")=0,
+       py::arg("slice_pitch")=0,
+       py::arg("host_buffer")=py::object()
+       ),
       py::return_value_policy<py::manage_new_object>());
 
   py::def("enqueue_read_image", enqueue_read_image,
-      (py::args("queue", "mem", "origin", "region", 
-                "row_pitch", "slice_pitch", "host_buffer"), 
+      (py::args("queue", "mem", "origin", "region", "host_buffer"), 
+       py::arg("row_pitch")=0,
+       py::arg("slice_pitch")=0,
        py::arg("wait_for")=py::object(),
        py::arg("is_blocking")=false),
       py::return_value_policy<py::manage_new_object>());
   py::def("enqueue_write_image", enqueue_write_image,
-      (py::args("queue", "mem", "origin", "region", 
-                "row_pitch", "slice_pitch", "host_buffer"), 
+      (py::args("queue", "mem", "origin", "region", "host_buffer"), 
+       py::arg("row_pitch")=0,
+       py::arg("slice_pitch")=0,
        py::arg("wait_for")=py::object(),
        py::arg("is_blocking")=false),
       py::return_value_policy<py::manage_new_object>());
@@ -584,6 +591,7 @@ BOOST_PYTHON_MODULE(_cl)
           py::return_self<>())
       .def(py::self == py::self)
       .def(py::self != py::self)
+      .def("all_kernels", create_kernels_in_program)
       ;
   }
 
@@ -608,7 +616,6 @@ BOOST_PYTHON_MODULE(_cl)
       ;
   }
 
-  DEF_SIMPLE_FUNCTION(create_kernels_in_program);
   py::def("enqueue_nd_range_kernel", enqueue_nd_range_kernel,
       (py::args("queue", "kernel"),
       py::arg("global_work_size"),
diff --git a/src/wrapper/wrap_cl.hpp b/src/wrapper/wrap_cl.hpp
index a8d94ef24f8805d6f7feb8c2faa2b68f4bb43d62..52a615936faedd45a958ea6f33183f49c1a9b2f8 100644
--- a/src/wrapper/wrap_cl.hpp
+++ b/src/wrapper/wrap_cl.hpp
@@ -19,7 +19,9 @@
 #include <iostream>
 #include <vector>
 #include <utility>
+#include <numeric>
 #include <boost/foreach.hpp>
+#include <boost/scoped_array.hpp>
 #include "wrap_helpers.hpp"
 #include "numpy_init.hpp"
 
@@ -400,13 +402,13 @@ namespace pyopencl
           case CL_DEVICE_VENDOR_ID: DEV_GET_INT_INF(cl_uint);
           case CL_DEVICE_MAX_COMPUTE_UNITS: DEV_GET_INT_INF(cl_uint);
           case CL_DEVICE_MAX_WORK_ITEM_DIMENSIONS: DEV_GET_INT_INF(cl_uint);
-          case CL_DEVICE_MAX_WORK_GROUP_SIZE: DEV_GET_INT_INF(cl_uint);
+          case CL_DEVICE_MAX_WORK_GROUP_SIZE: DEV_GET_INT_INF(size_t);
 
           case CL_DEVICE_MAX_WORK_ITEM_SIZES:
             {
               std::vector<size_t> result;
               PYOPENCL_GET_VEC_INFO(Device, m_device, param_name, result);
-              return py::list(result);
+              PYOPENCL_RETURN_VECTOR(size_t, result);
             }
 
           case CL_DEVICE_PREFERRED_VECTOR_WIDTH_CHAR: DEV_GET_INT_INF(cl_uint);
@@ -426,7 +428,7 @@ namespace pyopencl
           case CL_DEVICE_IMAGE3D_MAX_HEIGHT: DEV_GET_INT_INF(size_t);
           case CL_DEVICE_IMAGE3D_MAX_DEPTH: DEV_GET_INT_INF(size_t);
           case CL_DEVICE_IMAGE_SUPPORT: DEV_GET_INT_INF(cl_bool);
-          case CL_DEVICE_MAX_PARAMETER_SIZE: DEV_GET_INT_INF(cl_bool);
+          case CL_DEVICE_MAX_PARAMETER_SIZE: DEV_GET_INT_INF(size_t);
           case CL_DEVICE_MAX_SAMPLERS: DEV_GET_INT_INF(cl_uint);
           case CL_DEVICE_MEM_BASE_ADDR_ALIGN: DEV_GET_INT_INF(cl_uint);
           case CL_DEVICE_MIN_DATA_TYPE_ALIGN_SIZE: DEV_GET_INT_INF(cl_uint);
@@ -611,6 +613,10 @@ namespace pyopencl
                           handle_from_new_ptr(new platform(
                             reinterpret_cast<cl_platform_id>(result[i+1]))));
                     }
+
+                  case 0:
+                    break;
+
                   default:
                     throw error("Context.get_info", CL_INVALID_VALUE,
                         "unkown context_property key encountered");
@@ -1135,7 +1141,8 @@ namespace pyopencl
     PYOPENCL_CALL_GUARDED(clGetSupportedImageFormats, (
           ctx.data(), flags, image_type, 
           num_image_formats, &formats.front(), 0));
-    return py::list(formats);
+
+    PYOPENCL_RETURN_VECTOR(cl_image_format, formats);
   }
 
 
@@ -1215,8 +1222,8 @@ namespace pyopencl
       command_queue &cq,
       memory_object &mem,
       py::object py_origin, py::object py_region,
-      size_t row_pitch, size_t slice_pitch,
       py::object buffer,
+      size_t row_pitch, size_t slice_pitch,
       py::object py_wait_for,
       bool is_blocking
       )
@@ -1249,8 +1256,8 @@ namespace pyopencl
       command_queue &cq,
       memory_object &mem,
       py::object py_origin, py::object py_region,
-      size_t row_pitch, size_t slice_pitch,
       py::object buffer,
+      size_t row_pitch, size_t slice_pitch,
       py::object py_wait_for,
       bool is_blocking
       )
@@ -1629,19 +1636,24 @@ namespace pyopencl
             {
               std::vector<size_t> result;
               PYOPENCL_GET_VEC_INFO(Program, m_program, param_name, result);
-              return py::list(result);
+              PYOPENCL_RETURN_VECTOR(size_t, result);
             }
           case CL_PROGRAM_BINARIES:
             {
               std::vector<size_t> sizes;
-              PYOPENCL_GET_VEC_INFO(Program, m_program, param_name, sizes);
+              PYOPENCL_GET_VEC_INFO(Program, m_program, CL_PROGRAM_BINARY_SIZES, sizes);
+
+              size_t total_size = std::accumulate(sizes.begin(), sizes.end(), 0);
 
-              std::vector<std::vector<unsigned char> > result(sizes.size());
-              std::vector<unsigned char *> result_ptrs(sizes.size());
+              boost::scoped_array<unsigned char> result(
+                  new unsigned char[total_size]);
+              std::vector<unsigned char *> result_ptrs;
+
+              unsigned char *ptr = result.get();
               for (unsigned i = 0; i < sizes.size(); ++i)
               {
-                result[i].resize(sizes[i]);
-                result_ptrs.push_back(&result[i].front());
+                result_ptrs.push_back(ptr);
+                ptr += sizes[i];
               }
 
               PYOPENCL_CALL_GUARDED(clGetProgramInfo,
@@ -1649,10 +1661,14 @@ namespace pyopencl
                    &result_ptrs.front(), 0)); \
 
               py::list py_result;
+              ptr = result.get();
               for (unsigned i = 0; i < sizes.size(); ++i)
+              {
                 py_result.append(py::str(
-                      reinterpret_cast<char *>(&result[i].front()),
+                      reinterpret_cast<char *>(ptr),
                       sizes[i]));
+                ptr += sizes[i];
+              }
               return py_result;
             }
 
diff --git a/src/wrapper/wrap_helpers.hpp b/src/wrapper/wrap_helpers.hpp
index 3561235b825505b52daf6d3ae0b66c12878fd7e8..39a5012b195e0eb8a585b57141d2618019bd5aa1 100644
--- a/src/wrapper/wrap_helpers.hpp
+++ b/src/wrapper/wrap_helpers.hpp
@@ -108,6 +108,14 @@ namespace py = boost::python;
     else \
       throw std::runtime_error("unrecognized order specifier"); \
 
+#define PYOPENCL_RETURN_VECTOR(ITEMTYPE, NAME) \
+  { \
+    py::list pyopencl_result; \
+    BOOST_FOREACH(ITEMTYPE item, NAME) \
+      pyopencl_result.append(item); \
+    return pyopencl_result; \
+  }
+
 
 
 namespace
diff --git a/test/test_wrapper.py b/test/test_wrapper.py
new file mode 100644
index 0000000000000000000000000000000000000000..83c81b6eaf024162629443c21eb1276a77edf59b
--- /dev/null
+++ b/test/test_wrapper.py
@@ -0,0 +1,155 @@
+from __future__ import division
+import numpy
+import numpy.linalg as la
+
+
+
+
+def have_cl():
+    try:
+        import pyopencl
+        return True
+    except:
+        return False
+
+
+if have_cl():
+    import pyopencl as cl
+
+
+
+
+class TestCL:
+    disabled = not have_cl()
+
+    def test_get_info(self):
+        had_failures = [False]
+
+        QUIRKS = [
+                ("NVIDIA", [
+                    (cl.Device, cl.device_info.PLATFORM),
+                    ]),
+                ]
+
+        def find_quirk(quirk_list, cl_obj, info):
+            for quirk_plat_name, quirks in quirk_list:
+                if quirk_plat_name in platform.name:
+                    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):
+            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":
+                    info = getattr(info_cls, info_name)
+
+                    try:
+                        func(info)
+                    except:
+                        print "failed get_info", type(cl_obj), info_name
+
+                        if find_quirk(QUIRKS, cl_obj, info):
+                            print "(known quirk for %s)" % platform.name
+                        else:
+                            had_failures[0] = True
+                            raise
+
+        for platform in cl.get_platforms():
+            do_test(platform, cl.platform_info)
+
+            for device in platform.get_devices():
+                do_test(device, cl.device_info)
+
+                ctx = cl.Context([device])
+                do_test(ctx, cl.context_info)
+
+                props = 0
+                if (device.queue_properties
+                        & cl.command_queue_properties.PROFILING_ENABLE):
+                    profiling = True
+                    props = cl.command_queue_properties.PROFILING_ENABLE
+                queue = cl.CommandQueue(ctx,
+                        properties=props)
+                do_test(queue, cl.command_queue_info)
+
+                prg = cl.create_program_with_source(ctx, """
+                    __kernel void sum(__global float *a)
+                    { a[get_global_id(0)] *= 2; }
+                    """).build()
+                do_test(prg, cl.program_info)
+                do_test(prg, cl.program_build_info,
+                        lambda info: prg.get_build_info(device, info))
+
+                cl.unload_compiler() # just for the heck of it
+
+                mf = cl.mem_flags
+                n = 2000
+                a_buf = cl.create_buffer(ctx, 0, n*4)
+
+                do_test(a_buf, cl.mem_info)
+
+                kernel = prg.sum
+                do_test(kernel, cl.kernel_info)
+
+                evt = kernel(queue, (n,), a_buf)
+                do_test(evt, cl.event_info)
+
+                if profiling:
+                    evt.wait()
+                    do_test(evt, cl.profiling_info,
+                            lambda info: evt.get_profiling_info(info))
+
+                if device.image_support:
+                    if "NVIDIA" not in platform.name:
+                        # Samplers are crashy in Nvidia's "conformant" CL release
+                        smp = cl.Sampler(ctx, True,
+                                cl.addressing_mode.CLAMP,
+                                cl.filter_mode.NEAREST)
+                        do_test(smp, cl.sampler_info)
+
+                    img_format = cl.get_supported_image_formats(
+                            ctx, cl.mem_flags.READ_ONLY, cl.mem_object_type.IMAGE2D)[0]
+
+                    img = cl.create_image_2d(ctx, cl.mem_flags.READ_ONLY, img_format,
+                            128, 128, 0)
+                    do_test(img, cl.image_info,
+                            lambda info: img.get_image_info(info))
+                    img.image.depth
+
+        if had_failures[0]:
+            raise RuntimeError("get_info testing had errors")
+
+    def test_invalid_kernel_names_cause_failures(self):
+        for platform in cl.get_platforms():
+            for device in platform.get_devices():
+                ctx = cl.Context([device])
+                prg = cl.create_program_with_source(ctx, """
+                    __kernel void sum(__global float *a)
+                    { a[get_global_id(0)] *= 2; }
+                    """).build()
+
+                try:
+                    prg.sam
+                    raise RuntimeError("invalid kernel name did not cause error")
+                except AttributeError:
+                    pass
+
+
+
+
+
+
+
+if __name__ == "__main__":
+    # make sure that import failures get reported, instead of skipping the tests.
+    import pyopencl
+
+    from py.test.cmdline import main
+    main([__file__])