Esempio n. 1
0
def timing(d_x, d_y, d_z, num):
    t_sum = 0
    t2_sum = 0
    for repeat in range(NUM_REPEATS + 1):
        start = drv.Event()
        stop = drv.Event()
        start.record()

        for n in range(num):
            offset = n * N1
            add(numpy.uintp(int(d_x) + offset),
                numpy.uintp(int(d_y) + offset),
                numpy.uintp(int(d_z) + offset),
                numpy.int32(N1),
                grid=((N1 - 1) // block_size + 1, 1),
                block=(128, 1, 1),
                stream=streams[n])

        stop.record()
        stop.synchronize()
        elapsed_time = start.time_till(stop)
        if repeat > 0:
            t_sum += elapsed_time
            t2_sum += elapsed_time * elapsed_time
    t_ave = t_sum / NUM_REPEATS
    t_err = math.sqrt(t2_sum / NUM_REPEATS - t_ave * t_ave)
    print("Time = {:.6f} +- {:.6f} ms.".format(t_ave, t_err))
Esempio n. 2
0
 def get_sub_region(self, origin, size):
     assert origin + size <= self._size
     if self._base_buffer is None:
         base_buffer = self
     else:
         base_buffer = self._base_buffer
     new_ptr = numpy.uintp(self._ptr) + numpy.uintp(origin)
     return CuBufferAdapter(self._context_adapter,
                            self._device_idx,
                            size,
                            offset=self._offset + origin,
                            ptr=new_ptr,
                            base_buffer=base_buffer)
Esempio n. 3
0
def test_octile_graph_weighted():

    assert(OctileGraph.dtype.isalignedstruct)

    dfg = Graph(
        nodes={
            '!i': [0, 1, 2],
            'charge': [1, -1, 2],
            'conjugate': [False, True, True],
            'hybridization': [2, 3, 1]
        },
        edges={
            '!i': [0, 0],
            '!j': [1, 2],
            'length': [0.5, 1.0],
            '!w': [1.0, 2.0]
        },
        title='H2O')

    og = OctileGraph(dfg)
    assert(og.n_node == len(dfg.nodes))
    assert(og.p_octile != 0)
    assert(og.p_degree != 0)
    assert(og.p_node != 0)
    with pytest.raises(AttributeError):
        og.p_octile = np.uintp(0)
    with pytest.raises(AttributeError):
        og.p_degree = np.uintp(0)
    with pytest.raises(AttributeError):
        og.p_node = np.uintp(0)

    assert(og.node_t.isalignedstruct)
    for name in og.node_t.names:
        assert(name in dfg.nodes.columns)
    assert('charge' in og.node_t.names)
    assert('conjugate' in og.node_t.names)
    assert('hybridization' in og.node_t.names)

    assert(og.edge_t.isalignedstruct)
    assert(len(og.edge_t.names) == 2)
    assert('weight' in og.edge_t.names)
    assert('label' in og.edge_t.names)

    for name in og.edge_t['label'].names:
        assert(name in dfg.edges.columns)
    for name in dfg.edges.columns:
        if name in ['!i', '!j', '!w']:
            continue
        assert(name in og.edge_t['label'].names)
Esempio n. 4
0
def test_taintflag_distance_03():
    reset()

    inf = N.uintp(-1)
    assert t3a.distance(t1a, False, 1) == inf
    assert t3a.distance(t1b, False, 1) == inf
    assert t3a.distance(t1c, False, 1) == inf
    assert t3a.distance(t2a, False, 1) == 1
    assert t3a.distance(t2b, False, 1) == inf
    assert t3a.distance(t3a, False, 1) == 0
    assert t3a.distance(t3b, False, 1) == inf

    assert t3b.distance(t1a, False, 1) == inf
    assert t3b.distance(t1b, False, 1) == inf
    assert t3b.distance(t1c, False, 1) == inf
    assert t3b.distance(t2a, False, 1) == inf
    assert t3b.distance(t2b, False, 1) == 1
    assert t3b.distance(t3a, False, 1) == inf
    assert t3b.distance(t3b, False, 1) == 0

    assert t2a.distance(t1a, False, 1) == 1
    assert t2a.distance(t1b, False, 1) == 1
    assert t2a.distance(t1c, False, 1) == inf
    assert t2a.distance(t2a, False, 1) == 0
    assert t2a.distance(t2b, False, 1) == inf
    assert t2a.distance(t3a, False, 1) == inf
    assert t2a.distance(t3b, False, 1) == inf

    assert t2b.distance(t1a, False, 1) == inf
    assert t2b.distance(t1b, False, 1) == 1
    assert t2b.distance(t1c, False, 1) == 1
    assert t2b.distance(t2a, False, 1) == inf
    assert t2b.distance(t2b, False, 1) == 0
    assert t2b.distance(t3a, False, 1) == inf
    assert t2b.distance(t3b, False, 1) == inf
Esempio n. 5
0
 def test_numpy(self):
     """NumPy objects get serialized to readable JSON."""
     l = [
         np.float32(12.5),
         np.float64(2.0),
         np.float16(0.5),
         np.bool(True),
         np.bool(False),
         np.bool_(True),
         np.unicode_("hello"),
         np.byte(12),
         np.short(12),
         np.intc(-13),
         np.int_(0),
         np.longlong(100),
         np.intp(7),
         np.ubyte(12),
         np.ushort(12),
         np.uintc(13),
         np.ulonglong(100),
         np.uintp(7),
         np.int8(1),
         np.int16(3),
         np.int32(4),
         np.int64(5),
         np.uint8(1),
         np.uint16(3),
         np.uint32(4),
         np.uint64(5),
     ]
     l2 = [l, np.array([1, 2, 3])]
     roundtripped = loads(dumps(l2, cls=EliotJSONEncoder))
     self.assertEqual([l, [1, 2, 3]], roundtripped)
Esempio n. 6
0
 def test_numpy(self):
     """NumPy objects get serialized to readable JSON."""
     l = [
         np.float32(12.5),
         np.float64(2.0),
         np.float16(0.5),
         np.bool(True),
         np.bool(False),
         np.bool_(True),
         np.unicode_("hello"),
         np.byte(12),
         np.short(12),
         np.intc(-13),
         np.int_(0),
         np.longlong(100),
         np.intp(7),
         np.ubyte(12),
         np.ushort(12),
         np.uintc(13),
         np.ulonglong(100),
         np.uintp(7),
         np.int8(1),
         np.int16(3),
         np.int32(4),
         np.int64(5),
         np.uint8(1),
         np.uint16(3),
         np.uint32(4),
         np.uint64(5),
     ]
     l2 = [l, np.array([1, 2, 3])]
     roundtripped = loads(dumps(l2, cls=EliotJSONEncoder))
     self.assertEqual([l, [1, 2, 3]], roundtripped)
Esempio n. 7
0
    def _build_gpu_func(self):
        funcid = "query_{}".format(np.uintp(hash(self._expr)))

        # Preamble
        lines = ["from numba import cuda"]

        # Signature
        args = [_MASK_VAR] + self._colnames + self._refnames + [_SIZE_VAR]
        lines.append("def {}({}):".format(funcid, ",".join(args)))

        # Initialize the index variable and return immediately
        # when it exceeds the data size
        lines.append("    {} = cuda.grid(1)".format(_LOOP_VAR))
        lines.append("    if {} >= {}:".format(_LOOP_VAR, _SIZE_VAR))
        lines.append("        return")

        colnames = set(self._colnames)

        # Kernel body
        def _lift_to_array_access(m):
            name = m.group(0)
            if name in colnames:
                return "{}[{}]".format(name, _LOOP_VAR)
            else:
                return "{}".format(name)

        expr = re.sub(r"[_a-z]\w*", _lift_to_array_access, self._expr)
        lines.append("    {}[{}] = {}".format(_MASK_VAR, _LOOP_VAR, expr))

        # Evaluate the string to get the Python function
        body = "\n".join(lines)
        glbs = {}
        six.exec_(body, glbs)
        return glbs[funcid]
Esempio n. 8
0
def mpi_sendrecv_xla_encode_gpu(c, sendbuf, recvbuf, token, source, dest,
                                sendtag, recvtag, comm, status):
    from ..xla_bridge.mpi_xla_bridge import MPI_STATUS_IGNORE_ADDR
    from ..xla_bridge.mpi_xla_bridge_gpu import build_sendrecv_descriptor

    comm = unpack_hashable(comm)
    status = unpack_hashable(status)

    recv_shape = c.GetShape(recvbuf)
    recv_dtype = recv_shape.element_type()
    recv_dims = recv_shape.dimensions()

    # compute total number of elements in recv array
    recv_nitems = _np.prod(recv_dims, dtype=int)
    recv_dtype_handle = to_dtype_handle(recv_dtype)

    send_shape = c.GetShape(sendbuf)
    send_dtype = send_shape.element_type()
    send_dims = send_shape.dimensions()

    # compute total number of elements in send array
    send_nitems = _np.prod(send_dims, dtype=int)
    send_dtype_handle = to_dtype_handle(send_dtype)

    sh = xla_client.Shape.tuple_shape([
        xla_client.Shape.array_shape(recv_dtype, recv_dims),
        xla_client.Shape.token_shape(),
    ])

    if status is None:
        status_ptr = _np.uintp(MPI_STATUS_IGNORE_ADDR)
    else:
        status_ptr = to_mpi_ptr(status)

    descriptor = build_sendrecv_descriptor(
        send_nitems,
        dest,
        sendtag,
        send_dtype_handle,
        recv_nitems,
        source,
        recvtag,
        recv_dtype_handle,
        to_mpi_handle(comm),
        status_ptr,
    )

    return xla_client.ops.CustomCall(
        c,
        b"mpi_sendrecv",
        operands=(
            sendbuf,
            token,
        ),
        shape=sh,
        opaque=descriptor,
        has_side_effect=True,
    )
Esempio n. 9
0
def test_octile_graph_weighted():

    assert (OctileGraph.dtype.isalignedstruct)

    dfg = Graph(nodes={
        'index': [0, 1, 2],
        'columns': ['charge', 'conjugate', 'hybridization'],
        'data': [[1, False, 2], [-1, True, 3], [2, True, 1]]
    },
                edges={
                    'index': [0, 1],
                    'columns': ['!ij', 'length', '!w'],
                    'data': [[(0, 1), 0.5, 1.0], [(0, 2), 1.0, 2.0]]
                },
                title='H2O')

    og = OctileGraph(dfg)
    assert (og.n_node == len(dfg.nodes))
    assert (og.padded_size >= og.n_node and og.padded_size % 8 == 0)
    assert (og.n_octile == (og.padded_size // 8)**2)
    assert (og.p_octile != 0)
    assert (og.p_degree != 0)
    assert (og.p_node != 0)
    with pytest.raises(AttributeError):
        og.p_octile = np.uintp(0)
    with pytest.raises(AttributeError):
        og.p_degree = np.uintp(0)
    with pytest.raises(AttributeError):
        og.p_node = np.uintp(0)

    assert (og.node_type.isalignedstruct)
    for name in og.node_type.names:
        assert (name in dfg.nodes.columns)
    for name in dfg.nodes.columns:
        assert (name in og.node_type.names)

    assert (og.edge_type.isalignedstruct)
    assert (len(og.edge_type.names) == 2)
    assert ('weight' in og.edge_type.names)
    assert ('label' in og.edge_type.names)

    for name in og.edge_type['label'].names:
        assert (name in dfg.edges.columns)
    for name in dfg.edges.drop(['!ij', '!w'], axis=1).columns:
        assert (name in og.edge_type['label'].names)
Esempio n. 10
0
def to_mpi_handle(mpi_obj):
    """
    Returns the handle of the underlying C mpi object.

    Only defined for some MPI types (such as MPI_Comm), throws NotImplementedError
    otherwise.

    Note: This is not a pointer, but the actual C integer representation of the object.
    """
    return _np.uintp(_MPI._handleof(mpi_obj))
Esempio n. 11
0
def test_pycuda_times2():
  mod = SourceModule(TIMES2_SRC)

  times2 = mod.get_function("times2")

  src = np.ones((1000, 1), dtype=np.float32)
  dst = np.zeros_like(src)

  times2(
      drv.In(src), drv.Out(dst), np.uintp(len(src)),
      block=(32,1,1), grid=(32,1,1))
  assert (dst == 2*src).all()
 def __init__(self, array, struct_arr_ptr):
     self.data = cuda.to_device(array)
     self.shape, self.dtype = array.shape, array.dtype
     """
     numpy.getbuffer() needed due to lack of new-style buffer interface for
     scalar numpy arrays as of numpy version 1.9.1
     see: https://github.com/inducer/pycuda/pull/60
     """
     cuda.memcpy_htod(int(struct_arr_ptr),
                      numpy.getbuffer(numpy.int32(array.size)))
     cuda.memcpy_htod(int(struct_arr_ptr) + 8,
                      numpy.getbuffer(numpy.uintp(int(self.data))))
Esempio n. 13
0
 def process(self, time):
     # map the buffer
     dst_mapping = self.buffer_cu.map()
     ptr = np.uintp(dst_mapping.device_ptr_and_size()[0])
     self.color_them(
         ptr,
         np.int32(self.N),
         np.float32(time),
         grid=(self.N // 1024 + 1, 1, 1),
         block=(1024, 1, 1),
     )
     cuda_driver.Context.synchronize()
     dst_mapping.unmap()
Esempio n. 14
0
    def __init__(self, array, struct_arr_ptr):
        self.data = cuda.to_device(array)
        self.shape, self.dtype = array.shape, array.dtype
        """
        numpy.getbuffer() needed due to lack of new-style buffer interface for
        scalar numpy arrays as of numpy version 1.9.1

        see: https://github.com/inducer/pycuda/pull/60
        """
        cuda.memcpy_htod(int(struct_arr_ptr),
                         numpy.getbuffer(numpy.int32(array.size)))
        cuda.memcpy_htod(int(struct_arr_ptr) + 8,
                         numpy.getbuffer(numpy.uintp(int(self.data))))
Esempio n. 15
0
    def __init__(self, matrix, struct_ptr):

        self.data = cuda.to_device(matrix)
        self.shape, self.dtype = matrix.shape, matrix.dtype

        cuda.memcpy_htod(int(struct_ptr),
                         np.getbuffer(np.int32(np.shape(matrix)[0])))
        cuda.memcpy_htod(
            int(struct_ptr) + 4, np.getbuffer(np.int32(np.shape(matrix)[1])))
        cuda.memcpy_htod(
            int(struct_ptr) + 8, np.getbuffer(np.int32(matrix.strides[1])))
        cuda.memcpy_htod(
            int(struct_ptr) + 16, np.getbuffer(np.uintp(int(self.data))))
Esempio n. 16
0
def test_createNumpyDataset():
    ny, nx = 200, 100
    data = np.random.randn(ny, nx).astype(np.float32)

    dset = gdal_array.OpenArray(data)
    raster = isce.io.Raster(np.uintp(dset.this))

    assert (raster.width == nx)
    assert (raster.length == ny)
    assert (raster.datatype() == gdal.GDT_Float32)

    dset = None
    del raster
Esempio n. 17
0
    def _build_arg_buf(args):
        handlers = []

        arg_data = []
        format = ""
        for i, arg in enumerate(args):
            if isinstance(arg, np.number):
                arg_data.append(arg)
                format += arg.dtype.char
            elif isinstance(arg, (DeviceAllocation, PooledDeviceAllocation)):
                arg_data.append(int(arg))
                format += "P"
            elif isinstance(arg, ArgumentHandler):
                handlers.append(arg)
                arg_data.append(int(arg.get_device_alloc()))
                format += "P"
            elif isinstance(arg, np.ndarray):
                if isinstance(arg.base, ManagedAllocationOrStub):
                    arg_data.append(int(arg.base))
                    format += "P"
                else:
                    arg_data.append(arg)
                    format += "%ds" % arg.nbytes
            elif isinstance(arg, np.void):
                arg_data.append(_my_bytes(_memoryview(arg)))
                format += "%ds" % arg.itemsize
            else:
                cai = getattr(arg, "__cuda_array_interface__", None)
                if cai:
                    arg_data.append(cai["data"][0])
                    format += "P"
                    continue

                try:
                    gpudata = np.uintp(arg.gpudata)
                except AttributeError:
                    raise TypeError("invalid type on parameter #%d (0-based)" %
                                    i)
                else:
                    # for gpuarrays
                    arg_data.append(int(gpudata))
                    format += "P"

        from pycuda._pvt_struct import pack

        return handlers, pack(format, *arg_data)
Esempio n. 18
0
    def _build_cpu_func(self):
        funcid = "query_{}".format(np.uintp(hash(self._expr)))

        # Preamble
        lines = ["from numba import carray, types"]

        # Signature
        lines.append("def {}({}, {}):".format(funcid, _ARGS_VAR, _SIZE_VAR))

        # Unpack kernel arguments
        def _emit_assignment(var, idx, sz, ty):
            lines.append("    {} = carray({}[{}], {}, types.{})".format(
                var, _ARGS_VAR, idx, sz, ty))

        arg_idx = 1
        _emit_assignment(_MASK_VAR, 0, _SIZE_VAR, numba.types.int8)
        for name, type in zip(self._colnames, self._col_types):
            _emit_assignment(name, arg_idx, _SIZE_VAR, type.dtype)
            arg_idx += 1
        for name, type in zip(self._refnames, self._ref_types):
            _emit_assignment(name, arg_idx, 0, type.dtype)
            arg_idx += 1

        # Main loop
        lines.append("    for {} in range({}):".format(_LOOP_VAR, _SIZE_VAR))

        colnames = set(self._colnames)

        # Kernel body
        def _lift_to_array_access(m):
            name = m.group(0)
            if name in colnames:
                return "{}[{}]".format(name, _LOOP_VAR)
            else:
                return "{}[0]".format(name)

        expr = re.sub(r"[_a-z]\w*", _lift_to_array_access, self._expr)
        lines.append("        {}[{}] = {}".format(_MASK_VAR, _LOOP_VAR, expr))

        # Evaluate the string to get the Python function
        body = "\n".join(lines)
        glbs = {}
        six.exec_(body, glbs)
        return glbs[funcid]
Esempio n. 19
0
def mpi_recv_xla_encode_gpu(c, x, token, source, tag, comm, status):
    from ..xla_bridge.mpi_xla_bridge import MPI_STATUS_IGNORE_ADDR
    from ..xla_bridge.mpi_xla_bridge_gpu import build_recv_descriptor

    comm = unpack_hashable(comm)
    status = unpack_hashable(status)

    x_shape = c.GetShape(x)
    dtype = x_shape.element_type()
    dims = x_shape.dimensions()

    # compute total number of elements in array
    nitems = _np.prod(dims, dtype=int)
    dtype_handle = to_dtype_handle(dtype)

    sh = xla_client.Shape.tuple_shape([
        xla_client.Shape.array_shape(dtype, dims),
        xla_client.Shape.token_shape()
    ])

    if status is None:
        status_ptr = _np.uintp(MPI_STATUS_IGNORE_ADDR)
    else:
        status_ptr = to_mpi_ptr(status)

    descriptor = build_recv_descriptor(
        nitems,
        source,
        tag,
        to_mpi_handle(comm),
        dtype_handle,
        status_ptr,
    )

    out = xla_client.ops.CustomCall(
        c,
        b"mpi_recv",
        operands=(token, ),
        shape=sh,
        opaque=descriptor,
        has_side_effect=True,
    )

    return out
Esempio n. 20
0
    def to_array(self, dtype="float64"):

        cs = self.configuration_space
        array_sparse = super(DenseConfiguration, self).get_array()

        # initialize output array
        # TODO(LT): specify `dtype` flexibly
        array_dense = np.zeros(cs.size_dense, dtype=dtype)

        # process numerical hyperparameters
        if cs.nums:
            array_dense[cs.num_trg] = array_sparse[cs.num_src]

        # process categorical hyperparameters
        if cs.cats:
            cat_trg_offset = np.uintp(array_sparse[cs.cat_src])
            array_dense[cs.cat_trg + cat_trg_offset] = 1

        return array_dense
Esempio n. 21
0
def query_compile(expr):
    """Compile the query expression.

    This generates a CUDA Kernel for the query expression.  The kernel is
    cached for reuse.  All variable names, including both references to
    columns and references to variables in the calling environment, in the
    expression are passed as argument to the kernel. Thus, the kernel is
    reusable on any dataframe and in any environment.

    Parameters
    ----------
    expr : str
        The boolean expression

    Returns
    -------
    compiled: dict
        key "kernel" is the cuda kernel for the query.
        key "args" is a sequence of name of the arguments.
    """

    funcid = 'queryexpr_{:x}'.format(np.uintp(hash(expr)))
    # Load cache
    compiled = _cache.get(funcid)
    # Cache not found
    if compiled is None:
        info = query_parser(expr)
        fn = query_builder(info, funcid)
        args = info['args']
        # compile
        devicefn = cuda.jit(device=True)(fn)

        kernelid = 'kernel_{}'.format(funcid)
        kernel = _wrap_query_expr(kernelid, devicefn, args)

        compiled = info.copy()
        compiled['kernel'] = kernel
        # Store cache
        _cache[funcid] = compiled
    return compiled
Esempio n. 22
0
def test_taintflag_distance_05():
    reset()

    inf = N.uintp(-1)
    assert t3a.distance(t1a, True, 0) == inf
    assert t3a.distance(t1b, True, 0) == inf
    assert t3a.distance(t1c, True, 0) == inf
    assert t3a.distance(t2a, True, 0) == inf
    assert t3a.distance(t2b, True, 0) == inf
    assert t3a.distance(t3a, True, 0) == 0
    assert t3a.distance(t3b, True, 0) == inf

    assert t3b.distance(t1a, True, 0) == inf
    assert t3b.distance(t1b, True,
                        0) == inf  # distance throught passthrough flag
    assert t3b.distance(t1c, True,
                        0) == inf  # distance throught passthrough flag
    assert t3b.distance(t2a, True, 0) == inf
    assert t3b.distance(t2b, True,
                        0) == inf  # distance to passthrough flag (not through)
    assert t3b.distance(t3a, True, 0) == inf
    assert t3b.distance(t3b, True, 0) == 0

    assert t2a.distance(t1a, True, 0) == inf
    assert t2a.distance(t1b, True, 0) == inf
    assert t2a.distance(t1c, True, 0) == inf
    assert t2a.distance(t2a, True, 0) == 0
    assert t2a.distance(t2b, True, 0) == inf
    assert t2a.distance(t3a, True, 0) == inf
    assert t2a.distance(t3b, True, 0) == inf

    assert t2b.distance(t1a, True, 0) == inf
    assert t2b.distance(
        t1b, True, 0) == inf  # distance from passthrough flag (not through)
    assert t2b.distance(
        t1c, True, 0) == inf  # distance from passthrough flag (not through)
    assert t2b.distance(t2a, True, 0) == inf
    assert t2b.distance(t2b, True, 0) == 0
    assert t2b.distance(t3a, True, 0) == inf
    assert t2b.distance(t3b, True, 0) == inf
Esempio n. 23
0
def query_compile(expr):
    """Compile the query expression.

    This generates a CUDA Kernel for the query expression.  The kernel is
    cached for reuse.  All variable names, including both references to
    columns and references to variables in the calling environment, in the
    expression are passed as argument to the kernel. Thus, the kernel is
    reusable on any dataframe and in any environment.

    Parameters
    ----------
    expr : str
        The boolean expression

    Returns
    -------
    compiled: dict
        key "kernel" is the cuda kernel for the query.
        key "args" is a sequence of name of the arguments.
    """

    funcid = "queryexpr_{:x}".format(np.uintp(hash(expr)))
    # Load cache
    compiled = _cache.get(funcid)
    # Cache not found
    if compiled is None:
        info = query_parser(expr)
        fn = query_builder(info, funcid)
        args = info["args"]
        # compile
        devicefn = cuda.jit(device=True)(fn)

        kernelid = "kernel_{}".format(funcid)
        kernel = _wrap_query_expr(kernelid, devicefn, args)

        compiled = info.copy()
        compiled["kernel"] = kernel
        # Store cache
        _cache[funcid] = compiled
    return compiled
Esempio n. 24
0
class Matrix:

    # Float* + 4 integers
    mem_size = 8 + 2 * np.uintp(0).nbytes

    def __init__(self, matrix, struct_ptr):

        self.data = cuda.to_device(matrix)
        self.shape, self.dtype = matrix.shape, matrix.dtype

        cuda.memcpy_htod(int(struct_ptr),
                         np.getbuffer(np.int32(np.shape(matrix)[0])))
        cuda.memcpy_htod(
            int(struct_ptr) + 4, np.getbuffer(np.int32(np.shape(matrix)[1])))
        cuda.memcpy_htod(
            int(struct_ptr) + 8, np.getbuffer(np.int32(matrix.strides[1])))
        cuda.memcpy_htod(
            int(struct_ptr) + 16, np.getbuffer(np.uintp(int(self.data))))

    def __str__(self):
        return "Matrix: " + str(
            cuda.from_device(self.data, self.shape, self.dtype))
    def __call__(
        self,    
        columns_to_take, # A list 
        ):
        """
        Returns a Spearman rho value for each passed in function definition
        """
        assert(len(columns_to_take) == self._column_cardinality)
        # Convert the columns to a bit mask.... 
        column_mask = 0
        for column_index in columns_to_take:
            column_mask += 1 << column_index

        grid_size = (1, self._function_count // 32 + 1 )

        # Invoke...
        raw_gpu_func(
            np.uint16(self._row_count),
            self._gpu_ranks,

            self._gpu_left_binary_encoded,
            np.uint16(self._function_count),
            self._gpu_function_definitions,

            self._gpu_result_space,
            np.uintp(self._gpu_result_pitch),

            np.uint32( column_mask ),
            np.uint16( self._column_count),

            self._gpu_rho_space,

            block = (self._row_count, 32 ,1),
            grid  = grid_size 
        )
        
        drv.memcpy_dtoh(self._rho_space, self._gpu_rho_space)
        return self._rho_space.copy()
Esempio n. 26
0
    def _build_arg_buf(args):
        handlers = []

        arg_data = []
        format = ""
        for i, arg in enumerate(args):
            if isinstance(arg, np.number):
                arg_data.append(arg)
                format += arg.dtype.char
            elif isinstance(arg, (DeviceAllocation, PooledDeviceAllocation)):
                arg_data.append(int(arg))
                format += "P"
            elif isinstance(arg, ArgumentHandler):
                handlers.append(arg)
                arg_data.append(int(arg.get_device_alloc()))
                format += "P"
            elif isinstance(arg, np.ndarray):
                if isinstance(arg.base, ManagedAllocationOrStub):
                    arg_data.append(int(arg.base))
                    format += "P"
                else:
                    arg_data.append(arg)
                    format += "%ds" % arg.nbytes
            elif isinstance(arg, np.void):
                arg_data.append(_my_bytes(_memoryview(arg)))
                format += "%ds" % arg.itemsize
            else:
                try:
                    gpudata = np.uintp(arg.gpudata)
                except AttributeError:
                    raise TypeError("invalid type on parameter #%d (0-based)" % i)
                else:
                    # for gpuarrays
                    arg_data.append(int(gpudata))
                    format += "P"

        from pycuda._pvt_struct import pack
        return handlers, pack(format, *arg_data)
Esempio n. 27
0
def test_taintflag_distance_01():
    reset()

    printboth(*flags)

    inf = N.uintp(-1)
    assert t3a.distance(t1a) == 2
    assert t3a.distance(t1b) == 2
    assert t3a.distance(t1c) == inf
    assert t3a.distance(t2a) == 1
    assert t3a.distance(t2b) == inf
    assert t3a.distance(t3a) == 0
    assert t3a.distance(t3b) == inf

    assert t3b.distance(t1a) == inf
    assert t3b.distance(t1b) == 2
    assert t3b.distance(t1c) == 2
    assert t3b.distance(t2a) == inf
    assert t3b.distance(t2b) == 1
    assert t3b.distance(t3a) == inf
    assert t3b.distance(t3b) == 0

    assert t2a.distance(t1a) == 1
    assert t2a.distance(t1b) == 1
    assert t2a.distance(t1c) == inf
    assert t2a.distance(t2a) == 0
    assert t2a.distance(t2b) == inf
    assert t2a.distance(t3a) == inf
    assert t2a.distance(t3b) == inf

    assert t2b.distance(t1a) == inf
    assert t2b.distance(t1b) == 1
    assert t2b.distance(t1c) == 1
    assert t2b.distance(t2a) == inf
    assert t2b.distance(t2b) == 0
    assert t2b.distance(t3a) == inf
    assert t2b.distance(t3b) == inf
Esempio n. 28
0
reveal_type(np.string_())  # E: numpy.bytes_
reveal_type(np.object0())  # E: numpy.object_
reveal_type(np.void0(0))  # E: numpy.void

reveal_type(np.byte())  # E: {byte}
reveal_type(np.short())  # E: {short}
reveal_type(np.intc())  # E: {intc}
reveal_type(np.intp())  # E: {intp}
reveal_type(np.int0())  # E: {intp}
reveal_type(np.int_())  # E: {int_}
reveal_type(np.longlong())  # E: {longlong}

reveal_type(np.ubyte())  # E: {ubyte}
reveal_type(np.ushort())  # E: {ushort}
reveal_type(np.uintc())  # E: {uintc}
reveal_type(np.uintp())  # E: {uintp}
reveal_type(np.uint0())  # E: {uintp}
reveal_type(np.uint())  # E: {uint}
reveal_type(np.ulonglong())  # E: {ulonglong}

reveal_type(np.half())  # E: {half}
reveal_type(np.single())  # E: {single}
reveal_type(np.double())  # E: {double}
reveal_type(np.float_())  # E: {double}
reveal_type(np.longdouble())  # E: {longdouble}
reveal_type(np.longfloat())  # E: {longdouble}

reveal_type(np.csingle())  # E: {csingle}
reveal_type(np.singlecomplex())  # E: {csingle}
reveal_type(np.cdouble())  # E: {cdouble}
reveal_type(np.complex_())  # E: {cdouble}
Esempio n. 29
0
def gpu_ptr(data):
    return np.uintp(data.ptr)
Esempio n. 30
0
import numpy
from numpy import ma

from amuse.units import quantities
from amuse.units.quantities import VectorQuantity
from amuse.units.quantities import is_quantity
from amuse.datamodel.base import *
from amuse.support import exceptions
from amuse.rfi.async_request import FakeASyncRequest

try:
    if numpy.uintp().itemsize < 8: raise Exception()
    from amuse.datamodel.simple_hash import SimpleHash
    _SIMPLE_HASH_PRESENT_ = True
except BaseException as ex:
    _SIMPLE_HASH_PRESENT_ = False

_PREFER_SORTED_KEYS_ = True


class InMemoryAttributeStorage(AttributeStorage):
    def __init__(self):
        self.mapping_from_attribute_to_quantities = {}
        self.particle_keys = numpy.zeros(0)
        self.__version__ = 0

        self.sorted_keys = numpy.zeros(0, dtype=numpy.int32)
        self.sorted_indices = numpy.zeros(0)
        self.index_array = numpy.zeros(0, dtype=numpy.int32)
        self.keys_set = set([])
Esempio n. 31
0
        a = a + blockIdx.x;
        for (int idx = threadIdx.x; idx < a->datalen; idx += blockDim.x)
        {
            float *a_ptr = a->ptr;
            a_ptr[idx] *= 2;
        }
    }
    """)
func = mod.get_function("double_array")
func(struct_arr, block=(32, 1, 1), grid=(2, 1))

print("doubled arrays")
print(array1)
print(array2)

func(numpy.uintp(do2_ptr), block=(32, 1, 1), grid=(1, 1))
print("doubled second only")
print(array1)
print(array2)

if cuda.get_version() < (4, ):
    func.prepare("P", block=(32, 1, 1))
    func.prepared_call((2, 1), struct_arr)
else:
    func.prepare("P")
    block = (32, 1, 1)
    func.prepared_call((2, 1), block, struct_arr)


print("doubled again")
print(array1)
Esempio n. 32
0
np.string_()
np.object0()
np.void0(0)

np.byte()
np.short()
np.intc()
np.intp()
np.int0()
np.int_()
np.longlong()

np.ubyte()
np.ushort()
np.uintc()
np.uintp()
np.uint0()
np.uint()
np.ulonglong()

np.half()
np.single()
np.double()
np.float_()
np.longdouble()
np.longfloat()

np.csingle()
np.singlecomplex()
np.cdouble()
np.complex_()
Esempio n. 33
0
        a = a + blockIdx.x;
        for (int idx = threadIdx.x; idx < a->datalen; idx += blockDim.x)
        {
            float *a_ptr = a->ptr;
            a_ptr[idx] *= 2;
        }
    }
    """)
func = mod.get_function("double_array")
func(struct_arr, block=(32, 1, 1), grid=(2, 1))

print("doubled arrays")
print(array1)
print(array2)

func(numpy.uintp(do2_ptr), block=(32, 1, 1), grid=(1, 1))
print("doubled second only")
print(array1)
print(array2)

if cuda.get_version() < (4, ):
    func.prepare("P", block=(32, 1, 1))
    func.prepared_call((2, 1), struct_arr)
else:
    func.prepare("P")
    block = (32, 1, 1)
    func.prepared_call((2, 1), block, struct_arr)

print("doubled again")
print(array1)
print(array2)
Esempio n. 34
0
# Aliases
reveal_type(np.unicode_())  # E: numpy.str_
reveal_type(np.str0())  # E: numpy.str_

reveal_type(np.byte())  # E: numpy.signedinteger[numpy.typing._
reveal_type(np.short())  # E: numpy.signedinteger[numpy.typing._
reveal_type(np.intc())  # E: numpy.signedinteger[numpy.typing._
reveal_type(np.intp())  # E: numpy.signedinteger[numpy.typing._
reveal_type(np.int0())  # E: numpy.signedinteger[numpy.typing._
reveal_type(np.int_())  # E: numpy.signedinteger[numpy.typing._
reveal_type(np.longlong())  # E: numpy.signedinteger[numpy.typing._

reveal_type(np.ubyte())  # E: numpy.unsignedinteger[numpy.typing._
reveal_type(np.ushort())  # E: numpy.unsignedinteger[numpy.typing._
reveal_type(np.uintc())  # E: numpy.unsignedinteger[numpy.typing._
reveal_type(np.uintp())  # E: numpy.unsignedinteger[numpy.typing._
reveal_type(np.uint0())  # E: numpy.unsignedinteger[numpy.typing._
reveal_type(np.uint())  # E: numpy.unsignedinteger[numpy.typing._
reveal_type(np.ulonglong())  # E: numpy.unsignedinteger[numpy.typing._

reveal_type(np.half())  # E: numpy.floating[numpy.typing._
reveal_type(np.single())  # E: numpy.floating[numpy.typing._
reveal_type(np.double())  # E: numpy.floating[numpy.typing._
reveal_type(np.float_())  # E: numpy.floating[numpy.typing._
reveal_type(np.longdouble())  # E: numpy.floating[numpy.typing._
reveal_type(np.longfloat())  # E: numpy.floating[numpy.typing._

reveal_type(np.csingle())  # E: numpy.complexfloating[numpy.typing._
reveal_type(np.singlecomplex())  # E: numpy.complexfloating[numpy.typing._
reveal_type(np.cdouble())  # E: numpy.complexfloating[numpy.typing._
reveal_type(np.complex_())  # E: numpy.complexfloating[numpy.typing._
class TestNumpy:
    @staticmethod
    def test_get_numpy() -> None:
        """
        Test get_numpy when module is present
        """
        # Arrange

        # Act
        result = Numpy.get_numpy()

        # Assert
        assert result is np

    @staticmethod
    def test_get_numpy_missing(mocker: MockFixture) -> None:
        """
        Test get_numpy when module is missing
        """
        # Arrange
        mocker.patch.dict("sys.modules", {"numpy": None})

        # Act
        result = Numpy.get_numpy()

        # Assert
        assert result is None

    @staticmethod
    def test_get_numpy_missing_error(mocker: MockFixture) -> None:
        """
        Test get_numpy when module is missing raises error
        """
        # Arrange
        mocker.patch.dict("sys.modules", {"numpy": None})

        # Act / assert
        with pytest.raises(ImportError, match="foo"):
            Numpy.get_numpy(raise_error=True, custom_error_message="foo")

    @staticmethod
    @pytest.mark.parametrize("value, expected", [(np.array([1, 2, 3]), True),
                                                 ([1, 2, 3], False)])
    def test_is_numpy_object(value, expected) -> None:
        """
        Test is_numpy_object
        """
        # Arrange

        # Act
        result = Numpy.is_numpy_object(value)

        # Assert
        assert result == expected

    @staticmethod
    def test_get_numpy_primatives() -> None:
        """
        Test _get_numpy_primatives
        """
        # Arrange

        # Act
        result = Numpy._get_numpy_primatives(np)

        # Assert
        assert len(result) == 33  # Expected number of types
        for thing in result:
            assert "numpy" in getattr(thing, "__module__", "").split(
                ".")  # Check that type is from numpy
            assert type(thing) is type  # Check that each type is a type

    @staticmethod
    def test_encode_numpy_error():
        """ Test that the encode_numpy raises an error if no encoding is defined. """
        # Arrange
        value = "not a numpy"

        # Act & Assert
        with pytest.raises(NotImplementedError):
            Numpy.encode_numpy(value)

    @staticmethod
    @pytest.mark.parametrize(
        "value, expected",
        [
            # fmt: off
            (np.array([['balloons'], ['are'], ['awesome']
                       ]), [['balloons'], ['are'], ['awesome']]),
            (np.bool_(1), True),
            (np.byte(4), 4),
            (np.ubyte(4), 4),
            (np.short(4), 4),
            (np.ushort(4), 4),
            (np.intc(4), 4),
            (np.uintc(4), 4),
            (np.int_(4), 4),
            (np.uint(4), 4),
            (np.longlong(4), 4),
            (np.ulonglong(4), 4),
            (np.float16(4), 4),
            (np.single(4), 4),
            (np.double(4), 4),
            (np.longdouble(4), 4),
            (np.csingle(4), 4),
            (np.cdouble(4), 4),
            (np.clongdouble(4), 4),
            (np.int8(4), 4),
            (np.int16(4), 4),
            (np.int32(4), 4),
            (np.int64(4), 4),
            (np.uint8(4), 4),
            (np.uint16(4), 4),
            (np.uint32(4), 4),
            (np.uint64(4), 4),
            (np.intp(4), 4),
            (np.uintp(4), 4),
            (np.float32(4), 4),
            (np.float64(4), 4),
            (np.complex64(4), 4 + 0j),
            (np.complex128(4), 4 + 0j),
            (np.complex_(4), 4 + 0j),
            # fmt: on
        ],
    )
    def test_encode_numpy(value, expected) -> None:
        """
        Test encode_numpy
        """
        # Arrange

        # Act
        result = Numpy.encode_numpy(value)

        # Assert
        assert result == expected
Esempio n. 36
0
import numpy
from numpy import ma

from itertools import izip

from amuse.units import quantities
from amuse.units.quantities import VectorQuantity
from amuse.units.quantities import is_quantity
from amuse.datamodel.base import *
from amuse.support import exceptions

try:
  if numpy.uintp().itemsize<8: raise Exception()
  from amuse.datamodel.simple_hash import SimpleHash
  _SIMPLE_HASH_PRESENT_=True
except BaseException as ex:
  _SIMPLE_HASH_PRESENT_=False

_PREFER_SORTED_KEYS_=True

class InMemoryAttributeStorage(AttributeStorage):
    
    def __init__(self):
        self.mapping_from_attribute_to_quantities = {}
        self.particle_keys = numpy.zeros(0)
        self.__version__ = 0
        
        self.sorted_keys = numpy.zeros(0, dtype=numpy.int32)
        self.sorted_indices = numpy.zeros(0)
        self.index_array = numpy.zeros(0, dtype=numpy.int32)
        self.keys_set = set([])
Esempio n. 37
0
def test_table_typing_numpy():
    # Pulled from https://numpy.org/devdocs/user/basics.types.html

    # Numerics
    table = wandb.Table(columns=["A"], dtype=[NumberType])
    table.add_data(None)
    table.add_data(42)
    table.add_data(np.byte(1))
    table.add_data(np.short(42))
    table.add_data(np.ushort(42))
    table.add_data(np.intc(42))
    table.add_data(np.uintc(42))
    table.add_data(np.int_(42))
    table.add_data(np.uint(42))
    table.add_data(np.longlong(42))
    table.add_data(np.ulonglong(42))
    table.add_data(np.half(42))
    table.add_data(np.float16(42))
    table.add_data(np.single(42))
    table.add_data(np.double(42))
    table.add_data(np.longdouble(42))
    table.add_data(np.csingle(42))
    table.add_data(np.cdouble(42))
    table.add_data(np.clongdouble(42))
    table.add_data(np.int8(42))
    table.add_data(np.int16(42))
    table.add_data(np.int32(42))
    table.add_data(np.int64(42))
    table.add_data(np.uint8(42))
    table.add_data(np.uint16(42))
    table.add_data(np.uint32(42))
    table.add_data(np.uint64(42))
    table.add_data(np.intp(42))
    table.add_data(np.uintp(42))
    table.add_data(np.float32(42))
    table.add_data(np.float64(42))
    table.add_data(np.float_(42))
    table.add_data(np.complex64(42))
    table.add_data(np.complex128(42))
    table.add_data(np.complex_(42))

    # Booleans
    table = wandb.Table(columns=["A"], dtype=[BooleanType])
    table.add_data(None)
    table.add_data(True)
    table.add_data(False)
    table.add_data(np.bool_(True))

    # Array of Numerics
    table = wandb.Table(columns=["A"], dtype=[[NumberType]])
    table.add_data(None)
    table.add_data([42])
    table.add_data(np.array([1, 0], dtype=np.byte))
    table.add_data(np.array([42, 42], dtype=np.short))
    table.add_data(np.array([42, 42], dtype=np.ushort))
    table.add_data(np.array([42, 42], dtype=np.intc))
    table.add_data(np.array([42, 42], dtype=np.uintc))
    table.add_data(np.array([42, 42], dtype=np.int_))
    table.add_data(np.array([42, 42], dtype=np.uint))
    table.add_data(np.array([42, 42], dtype=np.longlong))
    table.add_data(np.array([42, 42], dtype=np.ulonglong))
    table.add_data(np.array([42, 42], dtype=np.half))
    table.add_data(np.array([42, 42], dtype=np.float16))
    table.add_data(np.array([42, 42], dtype=np.single))
    table.add_data(np.array([42, 42], dtype=np.double))
    table.add_data(np.array([42, 42], dtype=np.longdouble))
    table.add_data(np.array([42, 42], dtype=np.csingle))
    table.add_data(np.array([42, 42], dtype=np.cdouble))
    table.add_data(np.array([42, 42], dtype=np.clongdouble))
    table.add_data(np.array([42, 42], dtype=np.int8))
    table.add_data(np.array([42, 42], dtype=np.int16))
    table.add_data(np.array([42, 42], dtype=np.int32))
    table.add_data(np.array([42, 42], dtype=np.int64))
    table.add_data(np.array([42, 42], dtype=np.uint8))
    table.add_data(np.array([42, 42], dtype=np.uint16))
    table.add_data(np.array([42, 42], dtype=np.uint32))
    table.add_data(np.array([42, 42], dtype=np.uint64))
    table.add_data(np.array([42, 42], dtype=np.intp))
    table.add_data(np.array([42, 42], dtype=np.uintp))
    table.add_data(np.array([42, 42], dtype=np.float32))
    table.add_data(np.array([42, 42], dtype=np.float64))
    table.add_data(np.array([42, 42], dtype=np.float_))
    table.add_data(np.array([42, 42], dtype=np.complex64))
    table.add_data(np.array([42, 42], dtype=np.complex128))
    table.add_data(np.array([42, 42], dtype=np.complex_))

    # Array of Booleans
    table = wandb.Table(columns=["A"], dtype=[[BooleanType]])
    table.add_data(None)
    table.add_data([True])
    table.add_data([False])
    table.add_data(np.array([True, False], dtype=np.bool_))

    # Nested arrays
    table = wandb.Table(columns=["A"])
    table.add_data([[[[1, 2, 3]]]])
    table.add_data(np.array([[[[1, 2, 3]]]]))