kyupy/Demo.ipynb

{
 "cells": [
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "# Loading and Exploring Gate-Level Circuits"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Example of parsing the bench data format to make simple gate-level circuits."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 1,
   "metadata": {},
   "outputs": [],
   "source": [
    "from kyupy import bench\n",
    "\n",
    "# load a file\n",
    "b01 = bench.load('tests/b01.bench')\n",
    "\n",
    "# ... or specify the circuit as string \n",
    "mycircuit = bench.parse('input(a,b) output(o1,o2,o3) x=buf(a) o1=not(x) o2=buf(x) o3=buf(x)')"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Circuits are objects of the class `Circuit`."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 2,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "<Circuit 'tests/b01.bench' with 92 nodes, 130 lines, 4 ports>"
      ]
     },
     "execution_count": 2,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "b01"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 3,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "<Circuit with 10 nodes, 8 lines, 5 ports>"
      ]
     },
     "execution_count": 3,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "mycircuit"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Circuits are containers for two types of elements: nodes and lines.\n",
    "* A `Node` is a named entity in a circuit (e.g. a gate, a standard cell, a named signal, or a fan-out point) that has connections to other nodes.\n",
    "* A `Line` is a directional 1:1 connection between two Nodes.\n",
    "\n",
    "Use the `dump()` method to get a string representation of all nodes and their connections."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 4,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "None(0,1,2,3,4)\n",
      "0:__fork__\"a\"  >1\n",
      "1:__fork__\"b\"  \n",
      "2:__fork__\"o1\" <2 \n",
      "3:__fork__\"o2\" <4 \n",
      "4:__fork__\"o3\" <6 \n",
      "5:buf\"x\" <1 >0\n",
      "6:__fork__\"x\" <0 >3 >5 >7\n",
      "7:not\"o1\" <3 >2\n",
      "8:buf\"o2\" <5 >4\n",
      "9:buf\"o3\" <7 >6\n"
     ]
    }
   ],
   "source": [
    "print(mycircuit.dump())"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The first line of the dump starts with the circuit name (\"None\" for `mycircuit`), followed by the node-IDs of all the ports (inputs and outputs) of the circuit.\n",
    "\n",
    "Each of the following lines describes one node.\n",
    "Each node in the circuit has a unique ID, a type, a name, and line-connections. This information is given on each line in that order.\n",
    "\n",
    "A line in the circuit has a unique ID, a driver node and a receiver node. The connections in the dump show the direction (\">\" for output, \"<\" for input) and the line-ID. For example in `mycircuit`: Node-0 has one output connected to Line-1, and this Line-1 is connected to the input of Node-5.\n",
    "\n",
    "The `interface` is the list of nodes forming the ports (inputs and outputs):"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 5,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "[0:__fork__\"a\"  >1,\n",
       " 1:__fork__\"b\"  ,\n",
       " 2:__fork__\"o1\" <2 ,\n",
       " 3:__fork__\"o2\" <4 ,\n",
       " 4:__fork__\"o3\" <6 ]"
      ]
     },
     "execution_count": 5,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "mycircuit.interface"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "## Nodes\n",
    "\n",
    "There are two types of nodes: __forks__ and __cells__.\n",
    "\n",
    "Forks have the special type `__fork__` while cells can be of various types (`buf`, `not`, `and`, `nor`, etc.).\n",
    "Forks are used to label signals with names and to connect a one cell to multiple other cells (fan-out).\n",
    "The names among all forks and among all cells within a circuit are unique.\n",
    "Thus, a fork and a cell are allowed to share the same name.\n",
    "\n",
    "Nodes in circuits can be accessed by ID or by name."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 6,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "7:not\"o1\" <3 >2"
      ]
     },
     "execution_count": 6,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "mycircuit.nodes[7]"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 7,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "6:__fork__\"x\" <0 >3 >5 >7"
      ]
     },
     "execution_count": 7,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "mycircuit.forks['x']"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 8,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "5:buf\"x\" <1 >0"
      ]
     },
     "execution_count": 8,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "mycircuit.cells['x']"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Nodes have an `index` (the node ID), a `kind` (the type), a `name`, as well as `ins` (input pins) and `outs` (output pins)"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 9,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "(6, '__fork__', 'x', [0], [3, 5, 7])"
      ]
     },
     "execution_count": 9,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "n = mycircuit.nodes[6]\n",
    "n.index, n.kind, n.name, n.ins, n.outs"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The inputs and outputs of a node are lists containing `Line` objects."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 10,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "kyupy.circuit.Line"
      ]
     },
     "execution_count": 10,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "type(n.ins[0])"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "## Lines\n",
    "\n",
    "A line is a directional connection between one driving node (`driver`) and one reading node (`reader`).\n",
    "\n",
    "A line also knows to which node pins it is connected to: `driver_pin`, `reader_pin`."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 11,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "(5, 6:__fork__\"x\" <0 >3 >5 >7, 8:buf\"o2\" <5 >4, 1, 0)"
      ]
     },
     "execution_count": 11,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "l = mycircuit.nodes[6].outs[1]\n",
    "l.index, l.driver, l.reader, l.driver_pin, l.reader_pin"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "## Basic Analysis Examples\n",
    "### Cell type statistics"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 12,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "defaultdict(<class 'int'>, {'DFF': 5, 'AND': 1, 'NAND': 28, 'OR': 1, 'NOT': 10})\n"
     ]
    }
   ],
   "source": [
    "from collections import defaultdict\n",
    "\n",
    "counts = defaultdict(int)\n",
    "\n",
    "for n in b01.cells.values():\n",
    "    counts[n.kind] += 1\n",
    "\n",
    "print(counts)"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "### Tracing a scan chain"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 13,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "<Circuit 'b14' with 31715 nodes, 46891 lines, 91 ports>"
      ]
     },
     "execution_count": 13,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "from kyupy import verilog\n",
    "\n",
    "b14 = verilog.load('tests/b14.v.gz')\n",
    "b14"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 14,
   "metadata": {
    "scrolled": false
   },
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "chain length 287\n",
      "output test_so000\n",
      "NBUFFX8_RVT HFSBUF_36_76\n",
      "SDFFARX1_RVT wr_reg\n",
      "INVX4_RVT HFSINV_691_254\n",
      "INVX0_RVT HFSINV_2682_255\n",
      "SDFFARX1_RVT state_reg\n",
      "NBUFFX2_RVT ZBUF_55_inst_860\n",
      "SDFFARX1_RVT reg3_reg_28_\n",
      "SDFFARX1_RVT reg3_reg_27_\n",
      "SDFFARX1_RVT reg3_reg_26_\n",
      "...\n",
      "NBUFFX2_RVT ZBUF_1656_inst_2160\n",
      "SDFFARX1_RVT IR_reg_3_\n",
      "NBUFFX2_RVT ZBUF_85_inst_865\n",
      "SDFFARX1_RVT IR_reg_2_\n",
      "SDFFARX1_RVT IR_reg_1_\n",
      "SDFFARX1_RVT IR_reg_0_\n",
      "NBUFFX2_RVT ZBUF_17_inst_905\n",
      "NBUFFX4_RVT ZBUF_275_inst_906\n",
      "SDFFARX1_RVT B_reg\n",
      "input test_si000\n"
     ]
    }
   ],
   "source": [
    "chain = []\n",
    "cell = b14.cells['test_so000']\n",
    "chain.append(cell)\n",
    "while len(cell.ins) > 0:\n",
    "    cell = cell.ins[2 if 'SDFF' in cell.kind else 0].driver\n",
    "    if '__fork__' not in cell.kind:\n",
    "        chain.append(cell)\n",
    "        \n",
    "print('chain length', len(chain))\n",
    "for c in chain[:10]:\n",
    "    print(c.kind, c.name)\n",
    "print('...')\n",
    "for c in chain[-10:]:\n",
    "    print(c.kind, c.name)"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "### Determining Logic Depth of Nodes"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 15,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "<Circuit 'b14' with 31715 nodes, 46891 lines, 91 ports>"
      ]
     },
     "execution_count": 15,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "from kyupy import verilog\n",
    "\n",
    "b14 = verilog.load('tests/b14.v.gz')\n",
    "b14"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Calculate logic level (logic depth, distance from inputs or scan flip-flops) for each node in the circuit.\n",
    "Inputs and flip-flops themselves are level 0, **cells** driven by just inputs and flip-flops are level 1, and so on.\n",
    "**Fork** nodes have the same level as their driver, because they do not increase the logic depth."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 16,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "Maximum logic depth: 112\n"
     ]
    }
   ],
   "source": [
    "import numpy as np\n",
    "\n",
    "levels = np.zeros(len(b14.nodes), dtype='uint16')  # store level for each node.\n",
    "\n",
    "for cell in b14.topological_order():\n",
    "    if 'DFF' in cell.kind or 'input' == cell.kind:\n",
    "        levels[cell.index] = 0\n",
    "    elif '__fork__' == cell.kind:\n",
    "        levels[cell.index] = levels[cell.ins[0].driver.index]  # forks only have exactly one driver\n",
    "    else:\n",
    "        levels[cell.index] = max([levels[line.driver.index] for line in cell.ins]) + 1\n",
    "        \n",
    "print(f'Maximum logic depth: {np.max(levels)}')"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "List nodes with the highest depth and which nodes they are driving."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 17,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "depth: 112 node: __fork__     n2692  driving: SDFFARX1_RVT reg1_reg_29_  \n",
      "depth: 112 node: NAND2X0_RVT  U465   driving: __fork__     n2692         \n",
      "depth: 112 node: NAND2X0_RVT  U562   driving: __fork__     n2724         \n",
      "depth: 112 node: __fork__     n2724  driving: SDFFARX1_RVT reg0_reg_29_  \n",
      "depth: 112 node: __fork__     n2608  driving: SDFFARX1_RVT B_reg         \n",
      "depth: 112 node: NAND2X0_RVT  U170   driving: __fork__     n2608         \n",
      "depth: 111 node: NAND2X0_RVT  U5550  driving: __fork__     n2693         \n",
      "depth: 111 node: __fork__     n2660  driving: SDFFARX1_RVT reg2_reg_29_  \n",
      "depth: 111 node: AND2X2_RVT   U5560  driving: __fork__     n2660         \n",
      "depth: 111 node: __fork__     n2725  driving: SDFFARX1_RVT reg0_reg_28_  \n",
      "depth: 111 node: __fork__     n2693  driving: SDFFARX1_RVT reg1_reg_28_  \n",
      "depth: 111 node: __fork__     n362   driving: NAND2X0_RVT  U170          \n",
      "depth: 111 node: NAND2X0_RVT  U173   driving: __fork__     n362          \n",
      "depth: 111 node: __fork__     n600   driving: NAND2X0_RVT  U562          \n",
      "depth: 111 node: NAND2X0_RVT  U563   driving: __fork__     n600          \n",
      "depth: 111 node: NAND2X0_RVT  U565   driving: __fork__     n2725         \n",
      "depth: 111 node: NAND2X0_RVT  U466   driving: __fork__     n535          \n",
      "depth: 111 node: __fork__     n535   driving: NAND2X0_RVT  U465          \n",
      "depth: 110 node: __fork__     n4691  driving: AND2X2_RVT   U5560         \n",
      "depth: 110 node: NAND2X0_RVT  U5736  driving: __fork__     n790          \n"
     ]
    }
   ],
   "source": [
    "nodes_by_depth = np.argsort(levels)[::-1]\n",
    "\n",
    "for n_idx in nodes_by_depth[:20]:\n",
    "    n = b14.nodes[n_idx]\n",
    "    readers = ', '.join([f'{l.reader.kind:12s} {l.reader.name:14s}' for l in n.outs])\n",
    "    print(f'depth: {levels[n_idx]} node: {n.kind:12s} {n.name:6s} driving: {readers}')"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "# Working With Test Data and Logic Simulation"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Load a stuck-at fault test pattern set and expected fault-free responses from a STIL file."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 18,
   "metadata": {},
   "outputs": [],
   "source": [
    "from kyupy import verilog, stil\n",
    "from kyupy.logic import MVArray, BPArray\n",
    "from kyupy.logic_sim import LogicSim\n",
    "\n",
    "b14 = verilog.load('tests/b14.v.gz')\n",
    "s = stil.load('tests/b14.stuck.stil.gz')\n",
    "stuck_tests = s.tests(b14)\n",
    "stuck_responses = s.responses(b14)"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Tests and responses are instances of `MVArray`. Its `length` is the number of test vectors stored, its `width` is the number of values in a vector. By default, the stil parser returns 8-valued test vectors (`m=8`)."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 19,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "<MVArray length=1081 width=306 m=8 nbytes=330786>"
      ]
     },
     "execution_count": 19,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "stuck_tests"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The internal storage (an `ndarray` of `uint8`) is accessible via `data`. The first axis is the width, and the last axis goes along the test set."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 20,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "(306, 1081)"
      ]
     },
     "execution_count": 20,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "stuck_tests.data.shape"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The subscript accessor returns a string representation of the given test vector number. Possible values are '0', '1', '-', 'X', 'R', 'F', 'P', and 'N'."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 21,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "'P0--------------------11011111011001100111010101011101----------------------------------00-10111011010110011101110010111010111011101100010000110101111111011010101001010101010101010101001010110101001010101010101010110100000111111111111111011010100100101010010010101101010101001010100111010001010010000011100'"
      ]
     },
     "execution_count": 21,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "stuck_tests[1]"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 22,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "'--10000010010100010111--------------------------------0101010010101010110101001001010100--011111110011011111000111010101010111011101100010000110101111111011010101001010101010101010101001010110101001010101010101010110100000111111111111111011010100100101010010010101101010101001010101000111111111111111011101'"
      ]
     },
     "execution_count": 22,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "stuck_responses[1]"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The order of values in the vectors correspond to the circuit's interface followed by the scan flip-flops as they appear in `b14.cells`.\n",
    "The test data can be used directly in the simulators as they use the same ordering convention.\n",
    "\n",
    "The logic simulator uses bit-parallel storage of logic values, but our loaded test data uses one `uint8` per logic value.\n",
    "To convert the storage layout, we instanciate a `BPArray` for the input stimuli.\n",
    "The storage layout is more compact, but individual values cannot be easily accessed anymore."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 23,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "<BPArray length=1081 width=306 m=8 bytes=124848>"
      ]
     },
     "execution_count": 23,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "stuck_tests_bp = BPArray(stuck_tests)\n",
    "stuck_tests_bp"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 24,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "(306, 3, 136)"
      ]
     },
     "execution_count": 24,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "stuck_tests_bp.data.shape"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The following code performs a 8-valued logic simulation and stores the results in a new instance of `BPArray`.\n",
    "The packed array is unpacked into an `MVArray` for value access."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 25,
   "metadata": {},
   "outputs": [],
   "source": [
    "responses_bp = BPArray((stuck_tests_bp.width, len(stuck_tests_bp)))\n",
    "simulator = LogicSim(b14, sims=len(stuck_tests_bp))\n",
    "simulator.assign(stuck_tests_bp)\n",
    "simulator.propagate()\n",
    "simulator.capture(responses_bp)\n",
    "responses = MVArray(responses_bp)"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 26,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "'--10000010010100010111--------------------------------0101010010101010110101001001010100--011111110011011111000111010101010111011101100010000110101111111011010101001010101010101010101001010110101001010101010101010110100000111111111111111011010100100101010010010101101010101001010101000111111111111111011101'"
      ]
     },
     "execution_count": 26,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "responses[1]"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Compare simulation results to expected fault-free responses loaded from STIL. The first test fails, because it is a flush test while simulation implicitly assumes a standard test with a capture clock."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 27,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "mismatch for test pattern 0\n",
      "1080 of 1081 responses matched with simulator\n"
     ]
    }
   ],
   "source": [
    "matches = 0\n",
    "for i in range(len(responses)):\n",
    "    if responses[i] == stuck_responses[i]:\n",
    "        matches += 1\n",
    "    else:\n",
    "        print(f'mismatch for test pattern {i}')\n",
    "print(f'{matches} of {len(responses)} responses matched with simulator')"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Transition faults require test vector pairs for testing. These pairs are generated by `tests_loc`, assuming a launch-on-capture scheme (two functional clock cycles after scan-in)."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 28,
   "metadata": {},
   "outputs": [],
   "source": [
    "s = stil.load('tests/b14.transition.stil.gz')\n",
    "trans_tests = s.tests_loc(b14)\n",
    "trans_responses = s.responses(b14)"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 29,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "<MVArray length=1392 width=306 m=8 nbytes=425952>"
      ]
     },
     "execution_count": 29,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "trans_tests"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Possible values in the string representation are: '0', '1', '-', 'X', 'R' (rising transition), 'F' (falling transition), 'P' (positive pulse(s), 010), 'N' (negative pulse(s), 101)."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 30,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "'00--------------------RRRRRRFRRRRRRRRRRRFFRFRRRRRRRRRR----------------------------------00-00000001110100011111011010000000000000000011001001100101111110101110110001000100010100110111111101101000000111110011100010111000111R1111111111111111111111110001100100000110100000111010101110RFF00F000F0F00F00000FF01F'"
      ]
     },
     "execution_count": 30,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "trans_tests[1]"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "We validate these patterns with an 8-valued logic simulation"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 31,
   "metadata": {},
   "outputs": [],
   "source": [
    "trans_tests_bp = BPArray(trans_tests)\n",
    "responses_bp = BPArray((trans_tests_bp.width, len(trans_tests_bp)))\n",
    "simulator = LogicSim(b14, sims=len(trans_tests_bp))\n",
    "simulator.assign(trans_tests_bp)\n",
    "simulator.propagate()\n",
    "simulator.capture(responses_bp)\n",
    "responses = MVArray(responses_bp)"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 32,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "'--F00000F00F0F000F00FF--------------------------------01110101011100000101100000100110R0--0RRRRRRRNNNRNRPRNNNNNRFFRFRRRRRRR000000000011001001100101111110101110110001000100010100110111111101101000000111110011100010111000NNNNNNNNNNNNNNNNNNNNNNNNNNNNP0011001000001101000001110101011101RRRRRRRRRRRRRRRRRRRRP01R'"
      ]
     },
     "execution_count": 32,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "responses[1]"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The responses loaded from STIL only contain the final logic values. Use simple character replacements before comparing these. First test is again a flush test."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 33,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "mismatch for test pattern 0\n",
      "1391 of 1392 responses matched with simulator\n"
     ]
    }
   ],
   "source": [
    "matches = 0\n",
    "for i in range(len(responses)):\n",
    "    if trans_responses[i] == responses[i].replace('P','0').replace('N','1').replace('R','1').replace('F','0'):\n",
    "        matches += 1\n",
    "    else:\n",
    "        print(f'mismatch for test pattern {i}')\n",
    "print(f'{matches} of {len(responses)} responses matched with simulator')"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "# Working With Delay Information and Timing Simulation"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Delay data for gates and interconnect can be loaded from SDF files. In kyupy's timing simulators, delays are associated with the lines between nodes, not with the nodes themselves. Each line in the circuit has a rising delay, a falling delay, a negative pulse threshold, and a positive pulse threshold. "
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 34,
   "metadata": {},
   "outputs": [],
   "source": [
    "from kyupy import sdf\n",
    "from kyupy.saed import pin_index\n",
    "\n",
    "df = sdf.load('tests/b14.sdf.gz')\n",
    "lt = df.annotation(b14, pin_index, dataset=0, interconnect=False)"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The returned delay information is an `ndarray` with a set of delay values for each line in the circuit."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 35,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "(46891, 2, 2)"
      ]
     },
     "execution_count": 35,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "lt.shape"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Number of non-0 values loaded:"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 36,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "119676"
      ]
     },
     "execution_count": 36,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "(lt != 0).sum()"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The available timing simulators are `WaveSim` and `WaveSimCuda`.\n",
    "They work similarly to `LogicSim` in that they evaluate all cells in topological order.\n",
    "Instead of propagating a logic value, however, they propagate waveforms.\n",
    "\n",
    "`WaveSim` uses the numba just-in-time compiler for acceleration on CPU.\n",
    "It falls back to pure python if numba is not available. `WaveSimCuda` uses numba for GPU acceleration.\n",
    "If no CUDA card is available, it will fall back to pure python (not jit-compiled for CPU!).\n",
    "Pure python is too slow for most purposes.\n",
    "\n",
    "Both simulators operate data-parallel.\n",
    "The following instanciates a new engine for 32 independent timing simulations and each signal line in the circuit can carry at most 16 transitions. All simulators share the same circuit and the same line delay specification."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 37,
   "metadata": {},
   "outputs": [],
   "source": [
    "from kyupy.wave_sim import WaveSimCuda, TMAX\n",
    "import numpy as np\n",
    "\n",
    "wsim = WaveSimCuda(b14, lt, sims=32, wavecaps=16)"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "These are various memories allocated, with waveforms usually being the largest. "
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 38,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "Waveforms              : 93908.5 kiB\n",
      "State Allocation Table : 1113.4 kiB\n",
      "Circuit Timing         : 1484.5 kiB\n",
      "Circuit Netlist        : 732.7 kiB\n",
      "Capture Data           : 267.8 kiB\n",
      "Test Stimuli Data      : 3.6 kiB\n"
     ]
    }
   ],
   "source": [
    "def print_mem(name, arr):\n",
    "    print(f'{name}: {arr.size * arr.itemsize / 1024:.1f} kiB')\n",
    "    \n",
    "print_mem('Waveforms              ', wsim.state)\n",
    "print_mem('State Allocation Table ', wsim.sat)\n",
    "print_mem('Circuit Timing         ', wsim.timing)\n",
    "print_mem('Circuit Netlist        ', wsim.ops)\n",
    "print_mem('Capture Data           ', wsim.cdata)\n",
    "print_mem('Test Stimuli Data      ', wsim.tdata)"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "This is a typical simulation loop where the number of patterns is larger than the number of simulators available.\n",
    "We simulate `trans_tests_bp`.\n",
    "The timing simulator accepts 8-valued `BPArray`s, but it will return response (capture) data in a different format."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 39,
   "metadata": {},
   "outputs": [],
   "source": [
    "sims = 128  # len(trans_tests_bp)  # Feel free to simulate all tests if CUDA is set up correctly.\n",
    "\n",
    "cdata = np.zeros((len(wsim.interface), sims, 7))  # space to store all capture data\n",
    "\n",
    "for offset in range(0, sims, wsim.sims):\n",
    "    wsim.assign(trans_tests_bp, offset=offset)\n",
    "    wsim.propagate(sims=sims-offset)\n",
    "    wsim.capture(time=2.5, cdata=cdata, offset=offset)  # capture at time 2.5"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The capture data contains for each PI, PO, and scan flip-flop (axis 0), and each test (axis 1) seven values:\n",
    "0. Probability of capturing a 1 at the given capture time (same as next value, if no standard deviation given).\n",
    "1. A capture value decided by random sampling according to above probability.\n",
    "2. The final value (assume a very late capture time).\n",
    "3. True, if there was a premature capture (capture error), i.e. final value is different from captured value.\n",
    "4. Earliest arrival time. The time at which the output transitioned from its initial value.\n",
    "5. Latest stabilization time. The time at which the output transitioned to its final value.\n",
    "6. Overflow indicator. If non-zero, some signals in the input cone of this output had more transitions than specified in `wavecaps`. Some transitions have been discarded, the final values in the waveforms are still valid."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 40,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "(306, 128, 7)"
      ]
     },
     "execution_count": 40,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "cdata.shape"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "For validating against known logic values, take `cdata[...,1]`."
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 41,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "mismatch for test pattern 0\n",
      "127 of 128 responses matched with simulator\n"
     ]
    }
   ],
   "source": [
    "matches = 0\n",
    "\n",
    "for i in range(cdata.shape[1]):\n",
    "    response = ''.join('1' if x > 0.5 else '0' for x in cdata[..., i, 1])\n",
    "    if trans_responses[i].replace('-','0') == response:\n",
    "        matches += 1\n",
    "    else:\n",
    "        print(f'mismatch for test pattern {i}')\n",
    "print(f'{matches} of {cdata.shape[1]} responses matched with simulator')"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "The circuit delay is the maximum among all latest stabilization times:"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 42,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "2.0610005855560303"
      ]
     },
     "execution_count": 42,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "cdata[...,5].max()"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Check for overflows. If too many of them occur, increase `wavecaps` during engine instanciation:"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 43,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "0.0"
      ]
     },
     "execution_count": 43,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "cdata[...,6].sum()"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "Check for capture failures:"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 44,
   "metadata": {},
   "outputs": [
    {
     "data": {
      "text/plain": [
       "0.0"
      ]
     },
     "execution_count": 44,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "cdata[...,3].sum()"
   ]
  },
  {
   "cell_type": "markdown",
   "metadata": {},
   "source": [
    "# CUDA Support Notes\n",
    "\n",
    "Try this code to check if CUDA is set up correctly.\n",
    "\n",
    "If there is an error related to `nvvm`, you probably need to set up some environment variables:\n",
    "```\n",
    "%env LD_LIBRARY_PATH=/usr/local/cuda/lib64\n",
    "%env CUDA_HOME=/usr/local/cuda\n",
    "```\n",
    "If problems persist, refer to documentations for numba and cuda. "
   ]
  },
  {
   "cell_type": "code",
   "execution_count": 45,
   "metadata": {},
   "outputs": [
    {
     "name": "stdout",
     "output_type": "stream",
     "text": [
      "Found 1 CUDA devices\n",
      "id 0              b'TITAN V'                              [SUPPORTED]\n",
      "                      compute capability: 7.0\n",
      "                           pci device id: 0\n",
      "                              pci bus id: 2\n",
      "Summary:\n",
      "\t1/1 devices are supported\n"
     ]
    },
    {
     "data": {
      "text/plain": [
       "True"
      ]
     },
     "execution_count": 45,
     "metadata": {},
     "output_type": "execute_result"
    }
   ],
   "source": [
    "from numba import cuda\n",
    "\n",
    "cuda.detect()"
   ]
  },
  {
   "cell_type": "code",
   "execution_count": null,
   "metadata": {},
   "outputs": [],
   "source": []
  }
 ],
 "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.6.8"
  }
 },
 "nbformat": 4,
 "nbformat_minor": 2
}