{ "cells": [ { "cell_type": "markdown", "metadata": {}, "source": [ "# PyOpenCL: Arrays" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "## Setup code" ] }, { "cell_type": "code", "execution_count": 5, "metadata": { "collapsed": false }, "outputs": [], "source": [ "import pyopencl as cl\n", "import numpy as np\n", "import numpy.linalg as la" ] }, { "cell_type": "code", "execution_count": 6, "metadata": { "collapsed": false }, "outputs": [], "source": [ "a = np.random.rand(1024, 1024).astype(np.float32)" ] }, { "cell_type": "code", "execution_count": 7, "metadata": { "collapsed": false }, "outputs": [], "source": [ "ctx = cl.create_some_context()\n", "queue = cl.CommandQueue(ctx)" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "## Creating arrays" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "This notebook demonstrates working with PyOpenCL's arrays, which provide a friendlier (and more numpy-like) face on OpenCL's buffers. This is the module where they live:" ] }, { "cell_type": "code", "execution_count": 8, "metadata": { "collapsed": false }, "outputs": [], "source": [ "import pyopencl.array" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "Now transfer to a *device array*." ] }, { "cell_type": "code", "execution_count": 9, "metadata": { "collapsed": false }, "outputs": [], "source": [ "a_dev = cl.array.to_device(queue, a)" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "Works like a numpy array! (`shape`, `dtype`, `strides`)" ] }, { "cell_type": "code", "execution_count": 10, "metadata": { "collapsed": false }, "outputs": [ { "name": "stderr", "output_type": "stream", "text": [ "/home/andreas/src/env-3.5/lib/python3.5/site-packages/IPython/core/formatters.py:92: DeprecationWarning: DisplayFormatter._ipython_display_formatter_default is deprecated: use @default decorator instead.\n", " def _ipython_display_formatter_default(self):\n", "/home/andreas/src/env-3.5/lib/python3.5/site-packages/IPython/core/formatters.py:669: DeprecationWarning: PlainTextFormatter._singleton_printers_default is deprecated: use @default decorator instead.\n", " def _singleton_printers_default(self):\n" ] }, { "data": { "text/plain": [ "(1024, 1024)" ] }, "execution_count": 10, "metadata": {}, "output_type": "execute_result" } ], "source": [ "a_dev.shape" ] }, { "cell_type": "code", "execution_count": 11, "metadata": { "collapsed": false }, "outputs": [ { "data": { "text/plain": [ "dtype('float32')" ] }, "execution_count": 11, "metadata": {}, "output_type": "execute_result" } ], "source": [ "a_dev.dtype" ] }, { "cell_type": "code", "execution_count": 12, "metadata": { "collapsed": false }, "outputs": [ { "data": { "text/plain": [ "(4096, 4)" ] }, "execution_count": 12, "metadata": {}, "output_type": "execute_result" } ], "source": [ "a_dev.strides" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "## Working with arrays" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "**Goal:** Wanted to double all entries." ] }, { "cell_type": "code", "execution_count": 13, "metadata": { "collapsed": false }, "outputs": [], "source": [ "twice_a_dev = 2*a_dev" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "Easy to turn back into a `numpy` array." ] }, { "cell_type": "code", "execution_count": 14, "metadata": { "collapsed": false }, "outputs": [], "source": [ "twice_a = twice_a_dev.get()" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "Check!" ] }, { "cell_type": "code", "execution_count": 15, "metadata": { "collapsed": false }, "outputs": [ { "name": "stdout", "output_type": "stream", "text": [ "0.0\n" ] } ], "source": [ "print(la.norm(twice_a - 2*a))" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "Can just `print` the array, too." ] }, { "cell_type": "code", "execution_count": 16, "metadata": { "collapsed": false }, "outputs": [ { "name": "stdout", "output_type": "stream", "text": [ "[[ 0.77836961 0.28050834 0.6613102 ..., 1.91516626 0.61054963\n", " 1.56502569]\n", " [ 1.53493118 0.50901324 0.00827558 ..., 1.19049335 0.48224956\n", " 0.1369826 ]\n", " [ 0.50581717 1.01614654 0.32951528 ..., 1.3467046 1.45456564\n", " 1.40221345]\n", " ..., \n", " [ 1.47264338 1.11805999 1.55873811 ..., 1.87507105 1.08121443\n", " 1.99759185]\n", " [ 0.43069166 1.68386734 0.92028683 ..., 0.6744886 1.33184588\n", " 1.66233599]\n", " [ 1.22339284 1.3037529 0.3637082 ..., 0.82762784 0.23160546\n", " 1.58330226]]\n" ] } ], "source": [ "print(twice_a_dev)" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "----" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "Easy to evaluate arbitrary (elementwise) expressions." ] }, { "cell_type": "code", "execution_count": 13, "metadata": { "collapsed": false }, "outputs": [], "source": [ "import pyopencl.clmath" ] }, { "cell_type": "code", "execution_count": 14, "metadata": { "collapsed": false }, "outputs": [ { "data": { "text/plain": [ "array([[ 2.10573769, 2.50124764, 0.42696333, ..., 3.63703895,\n", " 1.92534614, 3.95568466],\n", " [ 4.62456608, 3.88678145, 1.25435662, ..., 1.32198358,\n", " -5.26086712, 4.57042027],\n", " [ 4.30697775, 2.99277115, 2.60830212, ..., 4.4082365 ,\n", " 2.17496896, 2.49961734],\n", " ..., \n", " [ 3.41792631, 0.50407267, -2.78950453, ..., 3.58545685,\n", " 4.49730206, 4.1767683 ],\n", " [ 3.05713558, 4.324893 , 4.29508495, ..., -9.8753109 ,\n", " 4.46689415, 3.88825035],\n", " [ 1.48695421, -2.0454402 , 3.78699446, ..., 2.17892647,\n", " 3.81082869, 3.16286278]], dtype=float32)" ] }, "execution_count": 14, "metadata": {}, "output_type": "execute_result" } ], "source": [ "cl.clmath.sin(a_dev)**2 - (1./a_dev) + 5" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "## Low-level Access\n", "\n", "Can still do everything manually though!" ] }, { "cell_type": "code", "execution_count": 15, "metadata": { "collapsed": false }, "outputs": [], "source": [ "prg = cl.Program(ctx, \"\"\"\n", " __kernel void twice(__global float *a)\n", " {\n", " int gid0 = get_global_id(0);\n", " int gid1 = get_global_id(1);\n", " int i = gid1 * 1024 + gid0;\n", " a[i] = 2*a[i];\n", " }\n", " \"\"\").build()\n", "twice = prg.twice" ] }, { "cell_type": "code", "execution_count": 16, "metadata": { "collapsed": false }, "outputs": [ { "data": { "text/plain": [ "" ] }, "execution_count": 16, "metadata": {}, "output_type": "execute_result" } ], "source": [ "twice(queue, a_dev.shape, None, a_dev.data)" ] }, { "cell_type": "code", "execution_count": 17, "metadata": { "collapsed": false }, "outputs": [ { "name": "stdout", "output_type": "stream", "text": [ "0.0 591.074\n" ] } ], "source": [ "print(la.norm(a_dev.get() - 2*a), la.norm(a))" ] }, { "cell_type": "markdown", "metadata": {}, "source": [ "But the hardcoded 1024 is ... inelegant. So fix that!\n", "\n", "(Also with arg `dtype setting`.)" ] } ], "metadata": { "kernelspec": { "display_name": "Python 3", "language": "python", "name": "python3" }, "language_info": { "codemirror_mode": { "name": "ipython", "version": 3 }, "file_extension": ".py", "mimetype": "text/x-python", "name": "python", "nbconvert_exporter": "python", "pygments_lexer": "ipython3", "version": "3.5.1+" } }, "nbformat": 4, "nbformat_minor": 0 }