Hello. I have found another error. I have been testing on ION (GeForce 9400M, CC 1.1, without double support) and one of the tests (test_mem_pool_with_arrays) gives error:
============================= test session starts
==============================
platform linux2 -- Python 2.6.7 -- pytest-1.3.4
test path 1: test_array.py
test_array.py ...........................................................F
=================================== FAILURES
===================================
test_mem_pool_with_arrays[ctx_factory=<context factory for
<pyopencl.Device 'ION' at 0x2f38f80>>]
ctx_factory = <pyopencl.tools.ContextFactory instance at 0x311c050>
@pytools.test.mark_test.opencl
def test_mem_pool_with_arrays(ctx_factory):
context = ctx_factory()
queue = cl.CommandQueue(context)
mem_pool = cl_tools.MemoryPool(cl_tools.CLAllocator(context))
> a_dev = cl_array.arange(queue, 2000, dtype=np.float64,
allocator=mem_pool)
test_array.py:631:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
def arange(*args, **kwargs):
"""Create an array filled with numbers spaced `step` apart,
starting from `start` and ending at `stop`.
For floating point arguments, the length of the result is
`ceil((stop - start)/step)`. This rule may result in the last
element of the result being greater than stop.
"""
if isinstance(args[0], cl.Context):
from warnings import warn
warn("Passing a context as first argument is deprecated. "
"This will be continue to be accepted througout "
"versions 2011.x of PyOpenCL.",
DeprecationWarning, 2)
args = args[1:]
> return _arange(*args, **kwargs)
/usr/lib/python2.6/dist-packages/pyopencl/array.py:813:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
queue = <pyopencl._cl.CommandQueue object at 0x3125578>
def _arange(queue, *args, **kwargs):
# argument processing
-----------------------------------------------------
# Yuck. Thanks, numpy developers. ;)
from pytools import Record
class Info(Record):
pass
explicit_dtype = False
inf = Info()
inf.start = None
inf.stop = None
inf.step = None
inf.dtype = None
inf.allocator = None
if isinstance(args[-1], np.dtype):
dtype = args[-1]
args = args[:-1]
explicit_dtype = True
argc = len(args)
if argc == 0:
raise ValueError, "stop argument required"
elif argc == 1:
inf.stop = args[0]
elif argc == 2:
inf.start = args[0]
inf.stop = args[1]
elif argc == 3:
inf.start = args[0]
inf.stop = args[1]
inf.step = args[2]
else:
raise ValueError, "too many arguments"
admissible_names = ["start", "stop", "step", "dtype",
"allocator"]
for k, v in kwargs.iteritems():
if k in admissible_names:
if getattr(inf, k) is None:
setattr(inf, k, v)
if k == "dtype":
explicit_dtype = True
else:
raise ValueError, "may not specify '%s' by position
and keyword" % k
else:
raise ValueError, "unexpected keyword argument '%s'" % k
if inf.start is None:
inf.start = 0
if inf.step is None:
inf.step = 1
if inf.dtype is None:
inf.dtype = np.array([inf.start, inf.stop, inf.step]).dtype
# actual functionality
----------------------------------------------------
dtype = np.dtype(inf.dtype)
start = dtype.type(inf.start)
step = dtype.type(inf.step)
stop = dtype.type(inf.stop)
if not explicit_dtype:
raise TypeError("arange requires dtype argument")
from math import ceil
size = int(ceil((stop-start)/step))
result = Array(queue, (size,), dtype, allocator=inf.allocator)
> _arange_knl(result, start, step, queue=queue)
/usr/lib/python2.6/dist-packages/pyopencl/array.py:790:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
def kernel_runner(*args, **kwargs):
repr_ary = args[0]
queue = kwargs.pop("queue", None) or repr_ary.queue
gs, ls = repr_ary.get_sizes(queue)
> knl = kernel_getter(*args)
/usr/lib/python2.6/dist-packages/pyopencl/array.py:147:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
result = array([ 0.00000000e+000, 5.26354425e-315, 0.00000000e
+000, ...,
1.76537186e-304, 1.76537882e-304, 1.76538751e-304])
start = 0.0, step = 1.0
@elwise_kernel_runner
def _arange_knl(result, start, step):
return elementwise.get_arange_kernel(
> result.context, result.dtype)
/usr/lib/python2.6/dist-packages/pyopencl/array.py:719:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
context = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
dtype = dtype('float64')
> ???
<string>:2:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
func = <function get_arange_kernel at 0x2dc8d70>
context = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
@decorator
def first_arg_dependent_memoize(func, context, *args):
"""Provides memoization for things that get created inside
a context, i.e. mainly programs and kernels. Assumes that
the first argument of the decorated function is the context.
"""
try:
ctx_dict = func._pyopencl_first_arg_dep_memoize_dic
except AttributeError:
# FIXME: This may keep contexts alive longer than desired.
# But I guess since the memory in them is freed, who cares.
ctx_dict = func._pyopencl_first_arg_dep_memoize_dic = {}
try:
return ctx_dict[context][args]
except KeyError:
first_arg_dependent_memoized_functions.append(func)
arg_dict = ctx_dict.setdefault(context, {})
> result = func(context, *args)
/usr/lib/python2.6/dist-packages/pyopencl/tools.py:69:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
context = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
dtype = dtype('float64')
@context_dependent_memoize
def get_arange_kernel(context, dtype):
return get_elwise_kernel(context,
"%(tp)s *z, %(tp)s start, %(tp)s step" % {
"tp": dtype_to_ctype(dtype),
},
"z[i] = start + i*step",
> name="arange")
/usr/lib/python2.6/dist-packages/pyopencl/elementwise.py:391:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
context = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
arguments = 'double *z, double start, double step'
operation = 'z[i] = start + i*step', name = 'arange', options = []
def get_elwise_kernel(context, arguments, operation,
name="elwise_kernel", options=[], **kwargs):
"""Return a L{pyopencl.Kernel} that performs the same scalar
operation
on one or several vectors.
"""
func, arguments = get_elwise_kernel_and_types(
context, arguments, operation,
> name=name, options=options, **kwargs)
/usr/lib/python2.6/dist-packages/pyopencl/elementwise.py:113:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
context = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
arguments = 'double *z, double start, double step'
operation = 'z[i] = start + i*step', name = 'arange', options = []
preamble = '#pragma OPENCL EXTENSION cl_khr_fp64: enable\n\n\n'
def get_elwise_kernel_and_types(context, arguments, operation,
name="elwise_kernel", options=[], preamble="", **kwargs):
if isinstance(arguments, str):
from pyopencl.tools import parse_c_arg
parsed_args = [parse_c_arg(arg) for arg in
arguments.split(",")]
else:
parsed_args = arguments
for arg in parsed_args:
if np.float64 == arg.dtype:
preamble = (
"#pragma OPENCL EXTENSION cl_khr_fp64: enable\n
\n\n"
+ preamble)
break
parsed_args.append(ScalarArg(np.uintp, "n"))
prg = get_elwise_program(
context, parsed_args, operation,
> name=name, options=options, preamble=preamble, **kwargs)
/usr/lib/python2.6/dist-packages/pyopencl/elementwise.py:91:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
context = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
arguments = [<[TypeError("not enough arguments for format string")
raised in repr()] VectorArg object at 0x3363fc8>, <[TypeError("... at
0x3363170>, <[TypeError("not enough arguments for format string") raised
in repr()] ScalarArg object at 0x3363050>]
operation = 'z[i] = start + i*step', name = 'arange', options = []
preamble = '#pragma OPENCL EXTENSION cl_khr_fp64: enable\n\n\n',
loop_prep = ''
after_loop = ''
def get_elwise_program(context, arguments, operation,
name="elwise_kernel", options=[],
preamble="", loop_prep="", after_loop=""):
from pyopencl import Program
source = ("""
%(preamble)s
__kernel void %(name)s(%(arguments)s)
{
unsigned lid = get_local_id(0);
unsigned gsize = get_global_size(0);
unsigned work_item_start =
get_local_size(0)*get_group_id(0);
unsigned i;
%(loop_prep)s;
for (i = work_item_start + lid; i < n; i += gsize)
{
%(operation)s;
}
%(after_loop)s;
}
""" % {
"arguments": ", ".join(arg.declarator() for arg in
arguments),
"operation": operation,
"name": name,
"preamble": preamble,
"loop_prep": loop_prep,
"after_loop": after_loop,
})
> return Program(context, source).build(options)
/usr/lib/python2.6/dist-packages/pyopencl/elementwise.py:69:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
self = <pyopencl.Program object at 0x349be10>
options = ['-I',
'/usr/lib/python2.6/dist-packages/pyopencl/../../../../include/pyopencl']
devices = None, cache_dir = None
def build(self, options=[], devices=None, cache_dir=None):
if isinstance(options, str):
options = [options]
options = options + ["-I", _find_pyopencl_include_path()]
if self._prg is not None:
self._prg._build(options, devices)
else:
from pyopencl.cache import
create_built_program_from_source_cached
self._prg = create_built_program_from_source_cached(
self._context, self._source, options, devices,
> cache_dir=cache_dir)
/usr/lib/python2.6/dist-packages/pyopencl/__init__.py:462:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
ctx = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
src = '\n #pragma OPENCL EXTENSION cl_khr_fp64: enable\n\n\n\n\n
__kernel void arange(__global double *z, doub...; i < n; i += gsize)\n
{\n z[i] = start + i*step;\n }\n\n ;
\n }\n '
options = ['-I',
'/usr/lib/python2.6/dist-packages/pyopencl/../../../../include/pyopencl']
devices = None, cache_dir = None
def create_built_program_from_source_cached(ctx, src, options=[],
devices=None,
cache_dir=None):
try:
if cache_dir != False:
prg, already_built =
_create_built_program_from_source_cached(
> ctx, src, options, devices, cache_dir)
/usr/lib/python2.6/dist-packages/pyopencl/cache.py:436:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
ctx = <pyopencl.Context at 0x339d3b0 on <pyopencl.Device 'ION' at
0x2f38f80>>
src = '\n #pragma OPENCL EXTENSION cl_khr_fp64: enable\n\n\n\n\n
__kernel void arange(__global double *z, doub...; i < n; i += gsize)\n
{\n z[i] = start + i*step;\n }\n\n ;
\n }\n '
options = ['-I',
'/usr/lib/python2.6/dist-packages/pyopencl/../../../../include/pyopencl']
devices = [<pyopencl.Device 'ION' at 0x2f38f80>]
cache_dir = '/tmp/pyopencl-compiler-cache-v2-uidtomus-py2.6.7.final.0'
def _create_built_program_from_source_cached(ctx, src, options,
devices, cache_dir):
include_path = ["."] + [
option[2:]
for option in options
if option.startswith("-I") or option.startswith("/I")]
if cache_dir is None:
from os.path import join
from tempfile import gettempdir
import getpass
cache_dir = join(gettempdir(),
"pyopencl-compiler-cache-v2-uid%s-py%s" % (
getpass.getuser(), ".".join(str(i) for i in
sys.version_info)))
# {{{ ensure cache directory exists
try:
os.mkdir(cache_dir)
except OSError, e:
from errno import EEXIST
if e.errno != EEXIST:
raise
# }}}
if devices is None:
devices = ctx.devices
cache_keys = [get_cache_key(device, options, src) for device in
devices]
binaries = []
to_be_built_indices = []
logs = []
for i, (device, cache_key) in enumerate(zip(devices,
cache_keys)):
cache_result = retrieve_from_cache(cache_dir, cache_key)
if cache_result is None:
to_be_built_indices.append(i)
binaries.append(None)
logs.append(None)
else:
binary, log = cache_result
binaries.append(binary)
logs.append(log)
message = (75*"="+"\n").join(
"Build on %s succeeded, but said:\n\n%s" % (dev, log)
for dev, log in zip(devices, logs)
if log is not None and log.strip())
if message:
from warnings import warn
warn("Build succeeded, but resulted in non-empty logs:
\n"+message)
# {{{ build on the build-needing devices, in one go
result = None
already_built = False
if to_be_built_indices:
prg = _cl._Program(ctx, src)
> prg.build(options, [devices[i] for i in
to_be_built_indices])
/usr/lib/python2.6/dist-packages/pyopencl/cache.py:361:
_ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _ _
_ _ _ _
self = <pyopencl._cl._Program object at 0x7ff712cc8520>
options =
'-I /usr/lib/python2.6/dist-packages/pyopencl/../../../../include/pyopencl'
devices = [<pyopencl.Device 'ION' at 0x2f38f80>]
def program_build(self, options=[], devices=None):
if isinstance(options, list):
options = " ".join(options)
err = None
try:
self._build(options=options, devices=devices)
except Exception, e:
from pytools import Record
class ErrorRecord(Record):
pass
what = e.what + "\n\n" + (75*"="+"\n").join(
"Build on %s:\n\n%s" % (dev, log)
for dev, log in self._get_build_logs())
code = e.code
routine = e.routine
err = _cl.RuntimeError(
ErrorRecord(
what=lambda : what,
code=lambda : code,
routine=lambda : routine))
if err is not None:
# Python 3.2 outputs the whole tree of currently active
exceptions
# This serves to remove one (redundant) level from that
nesting.
> raise err
E RuntimeError: clBuildProgram failed: invalid binary -
E
E Build on <pyopencl.Device 'ION' at 0x2f38f80>:
E
E ptxas application ptx input, line 52; error : Instruction
'cvt' requires SM 1.3 or higher, or map_f64_to_f32 directive
E ptxas application ptx input, line 53; error : Instruction
'ld' requires SM 1.3 or higher, or map_f64_to_f32 directive
E ptxas application ptx input, line 54; error : Instruction
'mul' requires SM 1.3 or higher, or map_f64_to_f32 directive
E ptxas application ptx input, line 55; error : Instruction
'ld' requires SM 1.3 or higher, or map_f64_to_f32 directive
E ptxas application ptx input, line 56; error : Instruction
'add' requires SM 1.3 or higher, or map_f64_to_f32 directive
E ptxas application ptx input, line 60; error : Instruction
'st' requires SM 1.3 or higher, or map_f64_to_f32 directive
E ptxas fatal : Ptx assembly aborted due to errors
E ptxas application ptx input, line 52; warning : Double is
not supported. Demoting to float
/usr/lib/python2.6/dist-packages/pyopencl/__init__.py:145: RuntimeError
===================== 1 failed, 59 passed in 97.97 seconds
=====================
After changing line 631 from
a_dev = cl_array.arange(queue, 2000, dtype=np.float64,
allocator=mem_pool)
to
a_dev = cl_array.arange(queue, 2000, dtype=np.float32,
allocator=mem_pool)
(i.e. to dtype=np.float32) everything passes.
Best regards.
--
Tomasz Rybak <[email protected]> GPG/PGP key ID: 2AD5 9860
Fingerprint A481 824E 7DD3 9C0E C40A 488E C654 FB33 2AD5 9860
http://member.acm.org/~tomaszrybak
signature.asc
Description: This is a digitally signed message part
_______________________________________________ PyOpenCL mailing list [email protected] http://lists.tiker.net/listinfo/pyopencl
