lab_vecAdd_PyCUDA.ipynb

{
 "cells": [
  {
   "cell_type": "code",
   "execution_count": 0,
   "metadata": {
    "colab": {},
    "colab_type": "code",
    "id": "ylUaUanV1wBZ"
   },
   "outputs": [],
   "source": [
    "!nvcc --version\n",
    "!gcc --version\n",
    "!g++ --version"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 0,
   "metadata": {
    "colab": {},
    "colab_type": "code",
    "id": "TmF7cxQv1yaX"
   },
   "outputs": [],
   "source": [
    "!sudo apt-get install python-numpy -y\n",
    "!sudo apt-get install build-essential python-dev python-setuptools libboost-python-dev libboost-thread-dev -y\n",
    "!apt-cache search pycuda\n",
    "!apt install libnvidia-compute-390 python3-pycuda"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 0,
   "metadata": {
    "colab": {},
    "colab_type": "code",
    "id": "2fqabneS5agF"
   },
   "outputs": [],
   "source": [
    "!apt-cache search pycuda\n",
    "!apt-get install python3-pycuda\n"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 51,
   "metadata": {
    "colab": {
     "base_uri": "https://localhost:8080/",
     "height": 90
    },
    "colab_type": "code",
    "id": "MRdlTrDV5Wg7",
    "outputId": "f370f1c9-6b80-4ad1-cd64-c8212fa4fea5"
   },
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "[1 0 0 ... 1 0 0]\n",
      "[0 1 0 ... 0 0 1]\n",
      "block: (256, 1, 1) grid: (4, 1)\n",
      "[1 1 0 ... 1 0 0]\n"
     ]
    }
   ],
   "source": [
    "import pycuda.driver as cuda\n",
    "import pycuda.autoinit\n",
    "from pycuda.compiler import SourceModule\n",
    "import numpy as np\n",
    "\n",
    "ELEM_NUMBER = 1024\n",
    "BLOCK_SIZE = 256\n",
    "\n",
    "A = np.random.randint(0, 2, dtype=np.int, size=ELEM_NUMBER)\n",
    "print (A)\n",
    "B = np.random.randint(0, 2, dtype=np.int, size=ELEM_NUMBER)\n",
    "print (B)\n",
    "\n",
    "A_gpu = cuda.mem_alloc(A.nbytes)\n",
    "B_gpu = cuda.mem_alloc(B.nbytes)\n",
    "C_gpu = cuda.mem_alloc(A.nbytes)\n",
    "\n",
    "\n",
    "cuda.memcpy_htod(A_gpu, A)\n",
    "cuda.memcpy_htod(B_gpu, B)\n",
    "\n",
    "mod = SourceModule(\"\"\"\n",
    "    __global__ void vecAdd(float *A, float *B, float *result) {\n",
    "\n",
    "      unsigned idx = blockIdx.x * blockDim.x + threadIdx.x;\n",
    "      result[idx] = A[idx] + B[idx];\n",
    "    }\n",
    "  \"\"\")\n",
    "\n",
    "func = mod.get_function(\"vecAdd\")\n",
    "\n",
    "\n",
    "gridDim = ( int((A.size + BLOCK_SIZE - 1)/BLOCK_SIZE),1)\n",
    "blockDim = (BLOCK_SIZE,1,1)\n",
    "print(\"block: \"+str(blockDim)+\" grid: \"+str(gridDim))\n",
    "\n",
    "func(A_gpu, B_gpu, C_gpu, block=blockDim, grid=gridDim)\n",
    "\n",
    "C = numpy.empty_like(A)\n",
    "cuda.memcpy_dtoh(C, C_gpu)\n",
    "print (C)\n"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 66,
   "metadata": {
    "colab": {
     "base_uri": "https://localhost:8080/",
     "height": 146
    },
    "colab_type": "code",
    "id": "zr83SKsZeFF1",
    "outputId": "23f199cf-fa18-4d11-a9b2-b6bbc6dde364"
   },
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "[0 0 1 ... 1 1 0]\n",
      "[0 1 1 ... 0 1 0]\n",
      "[0 0 1 ... 1 1 0]\n",
      "[0 1 1 ... 0 1 0]\n",
      "[0 0 0 ... 0 0 0]\n",
      "block: (256, 1, 1) grid: (4, 1)\n",
      "[0 1 2 ... 0 0 0]\n"
     ]
    }
   ],
   "source": [
    "import pycuda.gpuarray as gpuarray\n",
    "import pycuda.driver as cuda\n",
    "import pycuda.autoinit\n",
    "import numpy  as np\n",
    "\n",
    "NUMBER_ELEMENT = 1024\n",
    "BLOCK_SIZE = 256\n",
    "\n",
    "A = np.random.randint(0, 2, dtype=np.int, size=NUMBER_ELEMENT)\n",
    "print (A)\n",
    "\n",
    "B = np.random.randint(0, 2, dtype=np.int, size=NUMBER_ELEMENT)\n",
    "print (B)\n",
    "\n",
    "A_gpu = gpuarray.to_gpu(A)\n",
    "print (A_gpu.get())\n",
    "\n",
    "B_gpu = gpuarray.to_gpu(B)\n",
    "print (B_gpu.get())\n",
    "\n",
    "C_gpu = gpuarray.to_gpu(np.empty(NUMBER_ELEMENT).astype(np.int))\n",
    "print (C_gpu.get())\n",
    "\n",
    "mod = SourceModule(\"\"\"\n",
    "    __global__ void vecAdd(float *A, float *B, float *result) {\n",
    "\n",
    "      unsigned idx = blockIdx.x * blockDim.x + threadIdx.x;\n",
    "      result[idx] = A[idx] + B[idx];\n",
    "    }\n",
    "  \"\"\")\n",
    "\n",
    "gridDim = ( int((A.size + BLOCK_SIZE - 1)/BLOCK_SIZE),1)\n",
    "blockDim = (BLOCK_SIZE,1,1)\n",
    "print(\"block: \"+str(blockDim)+\" grid: \"+str(gridDim))\n",
    "\n",
    "func = mod.get_function(\"vecAdd\")\n",
    "func(A_gpu, B_gpu, C_gpu, block=blockDim, grid=gridDim)\n",
    "\n",
    "result = C_gpu.get()\n",
    "print (result)\n",
    "\n"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 75,
   "metadata": {
    "colab": {
     "base_uri": "https://localhost:8080/",
     "height": 72
    },
    "colab_type": "code",
    "id": "xPzjwaVL70Cq",
    "outputId": "14c017af-5a0b-4d7c-b9bd-78181e2a6d5a"
   },
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "Requirement already satisfied: numba in /usr/local/lib/python3.6/dist-packages (0.40.1)\n",
      "Requirement already satisfied: llvmlite>=0.25.0dev0 in /usr/local/lib/python3.6/dist-packages (from numba) (0.28.0)\n",
      "Requirement already satisfied: numpy in /usr/local/lib/python3.6/dist-packages (from numba) (1.14.6)\n"
     ]
    }
   ],
   "source": [
    "!pip3 install numba\n",
    "import os\n",
    "os.environ['NUMBAPRO_LIBDEVICE'] = \"/usr/lib/nvidia-cuda-toolkit/libdevice\"\n",
    "os.environ['NUMBAPRO_NVVM'] = \"/usr/lib/x86_64-linux-gnu/libnvvm.so\"\n"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 99,
   "metadata": {
    "colab": {
     "base_uri": "https://localhost:8080/",
     "height": 109
    },
    "colab_type": "code",
    "id": "9hxwTqxO4_Sk",
    "outputId": "58dfbb1e-2ca3-4ad2-9f2c-7e174c516ce3"
   },
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "[0 1 1 ... 1 1 0]\n",
      "[1 1 1 ... 1 1 0]\n",
      "[0 0 0 ... 0 0 0]\n",
      "Execution Time GPU is 185.06285095214844 ms\n",
      "[1 2 2 ... 2 2 0]\n"
     ]
    }
   ],
   "source": [
    "from numba import cuda\n",
    "import numpy  as np\n",
    "\n",
    "NUMBER_ELEMENT = 102400\n",
    "BLOCK_SIZE = 256\n",
    "\n",
    "@cuda.jit\n",
    "def vecADD(A,B,C):\n",
    "    # Thread id in a 1D block\n",
    "    idThreadX = cuda.threadIdx.x\n",
    "    # Block id in a 1D grid\n",
    "    idBlock = cuda.blockIdx.x\n",
    "    # Block width, i.e. number of threads per block\n",
    "    blockDim = cuda.blockDim.x\n",
    "    # Compute flattened index inside the array\n",
    "    idGlobal = idBlock * blockDim + idThreadX\n",
    "    if idGlobal < A.size:  # Check array boundaries\n",
    "        C[idGlobal] = A[idGlobal] + B[idGlobal]\n",
    "\n",
    "        \n",
    "A = np.random.randint(0, 2, dtype=np.int, size=NUMBER_ELEMENT)\n",
    "print (A)\n",
    "\n",
    "B = np.random.randint(0, 2, dtype=np.int, size=NUMBER_ELEMENT)\n",
    "print (B)\n",
    "\n",
    "C = np.empty(NUMBER_ELEMENT).astype(np.int)\n",
    "print (C)\n",
    "\n",
    "blockspergrid = (A.size + (BLOCK_SIZE - 1)) \n",
    "\n",
    "startEvent = cuda.event()\n",
    "stopEvent = cuda.event()\n",
    "\n",
    "startEvent.record()\n",
    "\n",
    "vecADD[blockspergrid, BLOCK_SIZE](A,B,C)     \n",
    "\n",
    "stopEvent.record()\n",
    "stopEvent.wait()\n",
    "stopEvent.synchronize()\n",
    "\n",
    "startEvent.elapsed_time(stopEvent)\n",
    "\n",
    "time = startEvent.elapsed_time(stopEvent)\n",
    "\n",
    "print(\"Execution Time GPU is \"+ str(time)+\" ms\" )\n",
    "print(C)"
   ]
  }
 ],
 "metadata": {
  "accelerator": "GPU",
  "colab": {
   "collapsed_sections": [],
   "name": "lab_vecAdd_PyCUDA.ipynb",
   "provenance": [],
   "version": "0.3.2"
  },
  "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.7.3"
  }
 },
 "nbformat": 4,
 "nbformat_minor": 1
}