disconnect unused clk from dff to prevent spurious sanity check fails

- sim: added LogicSim2V, LogicSim4V, LogicSim6V (faster and simpler API)
- sim: fixed model for: many-input gates in bench; dangling nodes; POs that drive further logic
- Circuit: dot output fix and better style, fixed equivalence check of lines and nodes.
- Techlib: Added KYUPY library with supported simulation primitives for debug and testing
- tests: fixed and added more for better coverage
- more type annotations
- updated project config and dependencies for better uv experience, newer lark, newer numpy

.gitignore

@@ -1,5 +1,14 @@
-**/__pycache__
-**/.ipynb_checkpoints
-**/.pytest_cache
-**/.DS_Store
-**/*.pyc
+__pycache__/
+.ipynb_checkpoints
+.pytest_cache
+.DS_Store
+.idea
+.vscode
+.venv
+*.py[oc]
+docs/_build
+build/
+dist/
+wheels/
+*.egg-info
+*nogit*

.python-version

@@ -0,0 +1 @@
+3.13

.readthedocs.yaml

@@ -0,0 +1,20 @@
+version: 2
+
+build:
+  os: "ubuntu-20.04"
+  tools:
+    python: "3.12"
+  jobs:
+    post_create_environment:
+      - python -m pip install sphinx_rtd_theme myst-parser
+
+sphinx:
+  fail_on_warning: true
+  configuration: docs/conf.py
+
+python:
+  # Install our python package before building the docs
+  install:
+    - method: pip
+      path: .
+

LICENSE.txt

@@ -1,6 +1,6 @@
 MIT License
 
-Copyright (c) 2020 s-holst
+Copyright (c) 2020-2025 Stefan Holst
 
 Permission is hereby granted, free of charge, to any person obtaining a copy
 of this software and associated documentation files (the "Software"), to deal

MANIFEST.in

@@ -1,5 +1,5 @@
-include *.ipynb
 include *.txt
+recursive-include examples *.ipynb
 recursive-include tests *.bench
 recursive-include tests *.gz
 recursive-include tests *.py

README.md

@@ -1,56 +1,32 @@
-KyuPy - Processing VLSI Circuits With Ease
-==========================================
+KyuPy - Pythonic Processing of VLSI Circuits
+============================================
 
-KyuPy is a python package for high-performance processing and analysis of
-non-hierarchical VLSI designs. Its purpose is to provide a rapid prototyping
-platform to aid and accelerate research in the fields of VLSI test, diagnosis
-and reliability. KyuPy is freely available under the MIT license.
+KyuPy is a Python package for processing and analysis of non-hierarchical gate-level VLSI designs.
+It contains fundamental building blocks for research software in the fields of VLSI test, diagnosis and reliability:
 
-
-Main Features
--------------
-
-* Partial [lark](https://github.com/lark-parser/lark) parsers for common files used with synthesized designs:
-  bench, gate-level verilog, standard delay format (SDF), standard test interface language (STIL)
-* Bit-parallel gate-level 2-, 4-, and 8-valued logic simulation
-* GPU-accelerated high-throughput gate-level timing simulation
-* High-performance through the use of [numpy](https://numpy.org) and [numba](https://numba.pydata.org)
+* Efficient data structures for gate-level circuits and related design data.
+* Partial [lark](https://github.com/lark-parser/lark) parsers for common design files like
+  bench, gate-level Verilog, standard delay format (SDF), standard test interface language (STIL), design exchange format (DEF).
+* Bit-parallel gate-level 2-, 4-, and 8-valued logic simulation.
+* GPU-accelerated high-throughput gate-level timing simulation.
+* High-performance through the use of [numpy](https://numpy.org) and [numba](https://numba.pydata.org).
 
 
 Getting Started
 ---------------
 
-KyuPy requires Python 3.6 or newer.
-Install the latest release by running:
-```commandline
-pip3 install --user kyupy
-```
-For best performance, ensure you have [numba](https://pypi.org/project/numba) installed:
-```commandline
-pip3 install --user numba
-```
-GPU/CUDA support may [require some additional setup](https://numba.pydata.org/numba-doc/latest/cuda/index.html).
-If CUDA or numba is not available, KyuPy will automatically fall back to slow, pure python execution.
+KyuPy is available in [PyPI](https://pypi.org/project/kyupy).
+It requires Python 3.10 or newer, [lark](https://pypi.org/project/lark), and [numpy](https://numpy.org).
+Although optional, [numba](https://numba.pydata.org) should be installed for best performance. [numba-cuda](https://nvidia.github.io/numba-cuda/) is required for GPU/CUDA acceleration.
+If numba is not available, KyuPy will automatically fall back to slow, pure Python execution.
 
-The Jupyter Notebook [UsageExamples.ipynb](https://github.com/s-holst/kyupy/blob/main/UsageExamples.ipynb) on GitHub
-contains some useful examples to get familiar with the API.
+The Jupyter Notebook [Introduction.ipynb](https://github.com/s-holst/kyupy/blob/main/examples/Introduction.ipynb) contains some useful examples to get familiar with the API.
 
 
 Development
 -----------
 
-To contribute to KyuPy or simply explore the source code, clone the KyuPy [repository](https://github.com/s-holst/kyupy) on GitHub.
-Within your local checkout, run:
-```commandline
-pip3 install --user -e .
-```
-to make the kyupy package available in your python environment.
-The source code comes with tests that can be run with:
-```
-pytest
-```
+To work with the latest pre-release source code, clone the [KyuPy GitHub repository](https://github.com/s-holst/kyupy).
 
-KyuPy depends on the following packages:
-* [lark-parser](https://pypi.org/project/lark-parser)
-* [numpy](https://pypi.org/project/numpy)
-* [numba](https://pypi.org/project/numba) (optional, required only for GPU/CUDA support)
+* Using ``pip``: Run ``pip install -e .`` within your local checkout to make the package available in your Python environment. The source code comes with tests that can be run with ``pytest``.
+* Using ``uv``: Run ``uv run pytest`` within your local checkout to get started.

docs/Makefile

@@ -0,0 +1,22 @@
+# pip install sphinx sphinx-rtd-theme myst-parser
+#
+# Minimal makefile for Sphinx documentation
+#
+
+# You can set these variables from the command line, and also
+# from the environment for the first two.
+SPHINXOPTS    ?=
+SPHINXBUILD   ?= sphinx-build
+SOURCEDIR     = .
+BUILDDIR      = _build
+
+# Put it first so that "make" without argument is like "make help".
+help:
+	@$(SPHINXBUILD) -M help "$(SOURCEDIR)" "$(BUILDDIR)" $(SPHINXOPTS) $(O)
+
+.PHONY: help Makefile
+
+# Catch-all target: route all unknown targets to Sphinx using the new
+# "make mode" option.  $(O) is meant as a shortcut for $(SPHINXOPTS).
+%: Makefile
+	@$(SPHINXBUILD) -M $@ "$(SOURCEDIR)" "$(BUILDDIR)" $(SPHINXOPTS) $(O)

docs/circuit.rst

@@ -0,0 +1,13 @@
+Circuit Graph - :mod:`kyupy.circuit`
+====================================
+
+.. automodule:: kyupy.circuit
+
+.. autoclass:: kyupy.circuit.Node
+   :members:
+
+.. autoclass:: kyupy.circuit.Line
+   :members:
+
+.. autoclass:: kyupy.circuit.Circuit
+   :members:

docs/conf.py

@@ -0,0 +1,65 @@
+# Configuration file for the Sphinx documentation builder.
+#
+# This file only contains a selection of the most common options. For a full
+# list see the documentation:
+# https://www.sphinx-doc.org/en/master/usage/configuration.html
+
+# -- Path setup --------------------------------------------------------------
+
+# If extensions (or modules to document with autodoc) are in another directory,
+# add these directories to sys.path here. If the directory is relative to the
+# documentation root, use os.path.abspath to make it absolute, like shown here.
+#
+import os
+import sys
+import sphinx_rtd_theme
+sys.path.insert(0, os.path.abspath('../src'))
+
+
+
+# -- Project information -----------------------------------------------------
+
+project = 'KyuPy'
+copyright = '2020-2026, Stefan Holst'
+author = 'Stefan Holst'
+
+# The full version, including alpha/beta/rc tags
+release = '0.0.7'
+
+
+# -- General configuration ---------------------------------------------------
+
+# Add any Sphinx extension module names here, as strings. They can be
+# extensions coming with Sphinx (named 'sphinx.ext.*') or your custom
+# ones.
+extensions = [
+    'sphinx.ext.autodoc',
+    'sphinx_rtd_theme',
+    'myst_parser',
+]
+
+# Add any paths that contain templates here, relative to this directory.
+templates_path = ['_templates']
+
+# List of patterns, relative to source directory, that match files and
+# directories to ignore when looking for source files.
+# This pattern also affects html_static_path and html_extra_path.
+exclude_patterns = ['_build', 'Thumbs.db', '.DS_Store']
+
+
+# -- Options for HTML output -------------------------------------------------
+
+# The theme to use for HTML and HTML Help pages.  See the documentation for
+# a list of builtin themes.
+#
+# html_theme = 'alabaster'
+html_theme = 'sphinx_rtd_theme'
+
+# Add any paths that contain custom static files (such as style sheets) here,
+# relative to this directory. They are copied after the builtin static files,
+# so a file named "default.css" will overwrite the builtin "default.css".
+#html_static_path = ['_static']
+
+autodoc_default_options = {
+    'member-order': 'bysource',
+}

docs/index.rst

@@ -0,0 +1,16 @@
+.. include:: ../README.md
+   :parser: myst_parser.sphinx_
+
+API Reference
+-------------
+
+.. toctree::
+   :maxdepth: 1
+
+   circuit
+   logic
+   techlib
+   parsers
+   simulators
+   miscellaneous
+

docs/logic.rst

@@ -0,0 +1,7 @@
+Multi-Valued Logic - :mod:`kyupy.logic`
+=======================================
+
+.. automodule:: kyupy.logic
+   :members:
+
+

docs/miscellaneous.rst

@@ -0,0 +1,6 @@
+Miscellaneous
+=============
+
+.. automodule:: kyupy
+   :members:
+

docs/parsers.rst

@@ -0,0 +1,51 @@
+Parsers
+=======
+
+KyuPy contains simple (and often incomplete) parsers for common file formats.
+These parsers are tailored to the most common use-cases to keep the grammars and the code-base as simple as possible.
+
+Each of the modules export a function ``parse()`` for parsing a string directly and a function
+``load()`` for loading a file. Files with a '.gz' extension are uncompressed on-the-fly.
+
+
+Verilog - :mod:`kyupy.verilog`
+------------------------------
+
+.. automodule:: kyupy.verilog
+   :members: parse, load
+
+
+Bench Format - :mod:`kyupy.bench`
+---------------------------------
+
+.. automodule:: kyupy.bench
+   :members: parse, load
+
+
+Standard Test Interface Language - :mod:`kyupy.stil`
+----------------------------------------------------
+
+.. automodule:: kyupy.stil
+   :members: parse, load
+
+.. autoclass:: kyupy.stil.StilFile
+   :members:
+
+
+Standard Delay Format - :mod:`kyupy.sdf`
+----------------------------------------
+
+.. automodule:: kyupy.sdf
+   :members: parse, load
+
+.. autoclass:: kyupy.sdf.DelayFile
+   :members:
+
+Design Exchange Format - :mod:`kyupy.def_file`
+----------------------------------------------
+
+.. automodule:: kyupy.def_file
+   :members: parse, load
+
+.. autoclass:: kyupy.def_file.DefFile
+   :members:

docs/simulators.rst

@@ -0,0 +1,37 @@
+Simulators
+==========
+
+KyuPy's simulators are optimized for cells with at most 4 inputs and 1 output.
+
+More complex cells must be mapped to simulation primitives first.
+
+
+Logic Simulation - :mod:`kyupy.logic_sim`
+-----------------------------------------
+
+.. automodule:: kyupy.logic_sim
+
+.. autoclass:: kyupy.logic_sim.LogicSim
+   :members:
+
+.. autoclass:: kyupy.logic_sim.LogicSim2V
+   :members:
+
+.. autoclass:: kyupy.logic_sim.LogicSim4V
+   :members:
+
+.. autoclass:: kyupy.logic_sim.LogicSim6V
+   :members:
+
+
+Timing Simulation - :mod:`kyupy.wave_sim`
+-----------------------------------------
+
+.. automodule:: kyupy.wave_sim
+   :members: TMAX, TMAX_OVL, TMIN
+
+.. autoclass:: kyupy.wave_sim.WaveSim
+   :members:
+
+.. autoclass:: kyupy.wave_sim.WaveSimCuda
+   :members:

docs/techlib.rst

@@ -0,0 +1,7 @@
+Technology Libraries
+====================
+
+.. automodule:: kyupy.techlib
+   :members:
+
+

pyproject.toml

@@ -0,0 +1,42 @@
+[project]
+name = "kyupy"
+version = "0.0.7"
+authors = [
+  { name="Stefan Holst", email="mail@s-holst.de" },
+]
+description = 'High-performance processing and analysis of non-hierarchical VLSI designs'
+readme = "README.md"
+requires-python = ">=3.12,<3.14"
+dependencies = [
+    "numpy>=1.17.0",
+    "lark>=1.3.0",
+]
+classifiers = [
+        "Development Status :: 3 - Alpha",
+        "Environment :: GPU :: NVIDIA CUDA",
+        "Intended Audience :: Science/Research",
+        "Topic :: Scientific/Engineering :: Electronic Design Automation (EDA)",
+        "License :: OSI Approved :: MIT License",
+        "Operating System :: OS Independent",
+        "Programming Language :: Python :: 3",
+]
+license = "MIT"
+license-files = ["LICENSE.txt"]
+
+[project.urls]
+homepage = "https://github.com/s-holst/kyupy"
+
+[build-system]
+requires = ["hatchling"]
+build-backend = "hatchling.build"
+
+[dependency-groups]
+dev = [
+    "myst-parser>=4.0.1",
+    "sphinx>=8.1.3",
+    "sphinx-rtd-theme>=3.0.2",
+    "pytest>=9.0.1",
+    "ipython>=9.14.1",
+    "jupyter>=1.1.1",
+]
+

setup.py

@@ -1,41 +0,0 @@
-from setuptools import setup, find_packages
-
-with open('README.md', 'r') as f:
-    long_description = f.read()
-
-setup(
-    name='kyupy',
-    version='0.0.1',
-    description='High-performance processing and analysis of non-hierarchical VLSI designs',
-    long_description=long_description,
-    long_description_content_type='text/markdown',
-    packages=find_packages(where='src'),
-    package_dir={'': 'src'},
-    url='https://github.com/s-holst/kyupy',
-    author='Stefan Holst',
-    author_email='mail@s-holst.de',
-    python_requires='>=3.6',
-    install_requires=[
-        'numpy>=1.15.0',
-        'lark-parser>=0.8.0'
-    ],
-    extras_requires={
-        'dev': [
-            'pytest>=6.1',
-        ],
-    },
-    classifiers=[
-        'Development Status :: 3 - Alpha',
-        'Environment :: GPU :: NVIDIA CUDA',
-        'Intended Audience :: Science/Research',
-        'Topic :: Scientific/Engineering :: Electronic Design Automation (EDA)',
-        'License :: OSI Approved :: MIT License',
-        'Operating System :: OS Independent',
-        'Programming Language :: Python :: 3',
-        'Programming Language :: Python :: 3 :: Only',
-        'Programming Language :: Python :: 3.6',
-        'Programming Language :: Python :: 3.7',
-        'Programming Language :: Python :: 3.8',
-        'Programming Language :: Python :: 3.9',
-    ],
-)

src/kyupy/__init__.py

@@ -1,35 +1,244 @@
-"""This package provides tools for high-performance processing and validation
-of non-hierarchical VLSI circuits to aid rapid prototyping of research code
-in the fields of VLSI test, diagnosis and reliability.
+"""The kyupy package itself contains a miscellaneous utility functions.
+
+In addition, it defines a ``numba`` and a ``cuda`` objects that point to the actual packages
+if they are available and otherwise point to mocks.
 """
 
 import time
+import sys
+from collections import defaultdict
 import importlib.util
+import gzip
+
+import numpy as np
+
+type NestedNumericList = list['int|NestedNumericList']
+type NestedStrIntDict = dict[str,'int|NestedStrIntDict']
+
+_pop_count_lut = np.asarray([bin(x).count('1') for x in range(256)])
+
+
+def cdiv(x, y):
+    return -(x // -y)
+
+
+def popcount(a):
+    """Returns the number of 1-bits in a given packed numpy array of type ``uint8``."""
+    return np.sum(_pop_count_lut[a])
+
+
+def readtext(file):
+    """Reads and returns the text in a given file. Transparently decompresses \\*.gz files."""
+    if hasattr(file, 'read'):
+        return file.read().decode()
+    if str(file).endswith('.gz'):
+        with gzip.open(file, 'rt') as f:
+            return f.read()
+    else:
+        with open(file, 'rt') as f:
+            return f.read()
+
+
+def hr_sci(value):
+    """Formats a value in a human-readible scientific notation."""
+    multiplier = 0
+    while abs(value) >= 1000:
+        value /= 1000
+        multiplier += 1
+    while abs(value) < 1:
+        value *= 1000
+        multiplier -= 1
+    return f'{value:.3f}{" kMGTPEafpnµm"[multiplier]}'
+
+
+def hr_bytes(nbytes):
+    """Formats a given number of bytes for human readability."""
+    multiplier = 0
+    while abs(nbytes) >= 1000:
+        nbytes /= 1024
+        multiplier += 1
+    return f'{nbytes:.1f}{["", "ki", "Mi", "Gi", "Ti", "Pi"][multiplier]}B'
+
+def eng(number):
+    """Formats a given number using engineering notation."""
+    exponent = 0
+    if abs(number) < 1:
+        while abs(number) >= 1000:
+            number *= 1000
+            exponent -= 3
+    else:
+        while abs(number) >= 1000:
+            number /= 1000
+            exponent += 3
+    return f'{number:.0f}' + (f'e{exponent}' if exponent != 0 else '')
+
+def hr_time(seconds):
+    """Formats a given time interval for human readability."""
+    s = ''
+    if seconds >= 86400:
+        d = seconds // 86400
+        seconds -= d * 86400
+        s += f'{int(d)}d'
+    if seconds >= 3600:
+        h = seconds // 3600
+        seconds -= h * 3600
+        s += f'{int(h)}h'
+    if seconds >= 60:
+        m = seconds // 60
+        seconds -= m * 60
+        if 'd' not in s:
+            s += f'{int(m)}m'
+    if 'h' not in s and 'd' not in s:
+        s += f'{int(seconds)}s'
+    return s
+
+
+def batchrange(nitems, maxsize):
+    """A simple generator that produces offsets and sizes for batch-loops."""
+    for offset in range(0, nitems, maxsize):
+        yield offset, min(nitems-offset, maxsize)
+
+
+class Timer:
+    def __init__(self, s=0): self.s = s
+    def __enter__(self): self.start_time = time.perf_counter(); return self
+    def __exit__(self, *args): self.s += time.perf_counter() - self.start_time
+    @property
+    def ms(self): return self.s*1e3
+    @property
+    def us(self): return self.s*1e6
+    def __repr__(self): return f'{self.s:.3f}'
+    def __add__(self, t):
+        return Timer(self.s + t.s)
+
+
+class Timers:
+    def __init__(self, t={}): self.timers: dict[str, Timer] = defaultdict(Timer) | t
+    def __getitem__(self, name): return self.timers[name]
+    def __repr__(self): return '{' + ', '.join([f'{k}: {v:.2f}' for k, v in self.dict().items()]) + '}'
+    def __add__(self, t):
+        tmr = Timers(self.timers)
+        for k, v in t.timers.items(): tmr.timers[k] += v
+        return tmr
+    def sum(self):
+        return sum([v.s for v in self.timers.values()])
+    def dict(self):
+        return dict(sorted([(k, v.s) for k, v in self.timers.items()], key=lambda item: -item[1]))
 
 
 class Log:
+    """A very simple logger that formats the messages with the number of seconds since
+    program start.
+    """
+
     def __init__(self):
         self.start = time.perf_counter()
-        self.logfile = None
+        self.logfile = sys.stdout
+        """When set to a file handle, log messages are written to it instead to standard output.
+        """
+        self.indent = 0
+        self._limit = -1
+        self.filtered = 0
+
+    def limit(self, log_limit):
+        class Limiter:
+            def __init__(self, l): self.l = l
+            def __enter__(self): self.l.start_limit(log_limit); return self
+            def __exit__(self, *args): self.l.stop_limit()
+        return Limiter(self)
+
+    def start_limit(self, limit):
+        self.filtered = 0
+        self._limit = limit
+
+    def stop_limit(self):
+        self._limit = -1
+        if self.filtered > 0:
+            self.info(f'{self.filtered} more messages (filtered).')
+            self.filtered = 0
+
+    def __getstate__(self):
+        return {'elapsed': time.perf_counter() - self.start}
+
+    def __setstate__(self, state):
+        self.logfile = sys.stdout
+        self.indent = 0
+        self._limit = -1
+        self.filtered = 0
+        self.start = time.perf_counter() - state['elapsed']
+
+    def write(self, s, indent=0):
+        self.logfile.write(' '*indent + s + '\n')
+        self.logfile.flush()
+
+    def li(self, item): self.write('- ' + str(item).replace('\n', '\n'+' '*(self.indent+1)), self.indent)
+    def lib(self): self.write('-', self.indent); self.indent += 1
+    def lin(self): self.write('-', self.indent-1)
+    def di(self, key, value): self.write(str(key) + ': ' + str(value).replace('\n', '\n'+' '*(self.indent+1)), self.indent)
+    def dib(self, key): self.write(str(key) + ':', self.indent); self.indent += 1
+    def din(self, key): self.write(str(key) + ':', self.indent-1)
+    def ie(self, n=1): self.indent -= n
 
     def log(self, level, message):
+        if self._limit == 0:
+            self.filtered += 1
+            return
         t = time.perf_counter() - self.start
-        if self.logfile is None:
-            print(f'{t:011.3f} {level} {message}')
-        else:
-            self.logfile.write(f'{t:011.3f} {level} {message}\n')
-            self.logfile.flush()
+        self.logfile.write(f'# {t:011.3f} {level} {message}\n')
+        #self.logfile.flush()
+        self._limit -= 1
 
-    def info(self, message): self.log('-', message)
+    def info(self, message):
+        """Log an informational message."""
+        self.log('-', message)
 
-    def warn(self, message): self.log('W', message)
+    def warn(self, message):
+        """Log a warning message."""
+        self.log('W', message)
 
-    def error(self, message): self.log('E', message)
+    def error(self, message):
+        """Log an error message."""
+        self.log('E', message)
+
+    def progress(self):
+        class Progress:
+            def __init__(self, l): self.l = l
+            def __enter__(self):
+                self.start_time = time.perf_counter()
+                self.lastlog_time = self.start_time
+                self.log_interval = 5
+                return self
+            def update(self, completion: float, msg: str = ''):
+                assert completion > 0 and completion <= 1
+                current_time = time.perf_counter()
+                if current_time > self.lastlog_time + self.log_interval:
+                    elapsed_time = current_time - self.start_time
+                    total_time = elapsed_time / completion
+                    rem_time = total_time - elapsed_time
+                    self.l.log(':', f'{msg} - {completion*100:.0f}% done {hr_time(elapsed_time)} elapsed {hr_time(rem_time)} remaining')
+                    self.log_interval = min(600, int(self.log_interval*1.5))
+                    self.lastlog_time = current_time
+            def __exit__(self, *args): pass
+        return Progress(self)
+
+    def range(self, *args):
+        """A generator that operates just like the ``range()`` built-in, and also occasionally logs the progress
+        and compute time estimates."""
+        elems = len(range(*args))
+        with self.progress() as p:
+            for elem, i in enumerate(range(*args)):
+                yield i
+                p.update((elem + 1) / elems)
 
 
 log = Log()
+"""The standard logger instance."""
 
 
+#
+# Code below mocks basic numba and cuda functions for pure-python fallback.
+#
+
 class MockNumba:
     @staticmethod
     def njit(func):
@@ -44,35 +253,35 @@ class MockCuda:
         self.x = 0
         self.y = 0
 
-    def jit(self, device=False):
+    def jit(self, func=None, device=False):
         _ = device  # silence "not used" warning
         outer = self
 
         def make_launcher(func):
-            class Launcher(object):
+            class Launcher:
                 def __init__(self, funcc):
                     self.func = funcc
 
                 def __call__(self, *args, **kwargs):
-                    # print(f'device func call {self.func.__name__}')
                     return self.func(*args, **kwargs)
 
                 def __getitem__(self, item):
                     grid_dim, block_dim = item
-                    # print(f'kernel call {self.func.__name__} grid_dim:{grid_dim} block_dim:{block_dim}')
 
                     def inner(*args, **kwargs):
                         for grid_x in range(grid_dim[0]):
                             for grid_y in range(grid_dim[1]):
                                 for block_x in range(block_dim[0]):
                                     for block_y in range(block_dim[1]):
-                                        outer.x = grid_x * block_dim[0] + block_x
-                                        outer.y = grid_y * block_dim[1] + block_y
+                                        outer.x = grid_x * \
+                                            block_dim[0] + block_x
+                                        outer.y = grid_y * \
+                                            block_dim[1] + block_y
                                         self.func(*args, **kwargs)
                     return inner
             return Launcher(func)
 
-        return make_launcher
+        return make_launcher(func) if func else make_launcher
 
     @staticmethod
     def to_device(array, to=None):
@@ -97,12 +306,16 @@ if importlib.util.find_spec('numba') is not None:
         list(numba.cuda.gpus)
         from numba import cuda
     except CudaSupportError:
-        log.warn('Cuda unavailable. Falling back to pure python')
+        log.warn('Cuda unavailable. Falling back to pure Python.')
         cuda = MockCuda()
 else:
     numba = MockNumba()
+    """If Numba is available on the system, it is the actual ``numba`` package.
+    Otherwise, it simply defines an ``njit`` decorator that does nothing.
+    """
     cuda = MockCuda()
-    log.warn('Numba unavailable. Falling back to pure python')
-
-
-
+    """If Numba is installed and Cuda GPUs are available, it is the actual ``numba.cuda`` package.
+    Otherwise, it is an object that defines basic methods and decorators so that cuda-code can still
+    run in the Python interpreter.
+    """
+    log.warn('Numba unavailable. Falling back to pure Python.')

src/kyupy/atalanta.py

@@ -0,0 +1,64 @@
+"""A parser and dumper for Atalanta's plain-text test pattern format (*.test).
+
+A *.test file contains one test pattern per line in the form ``<index>: <bits>``,
+where ``<bits>`` is a string of ``'0'`` and ``'1'`` characters giving the values of
+the primary inputs (in circuit order). Lines starting with ``'*'`` are comments.
+
+This parser also accepts files without ``<index>:`` and patterns elements in 8-valued
+logic such as ``'R'``, ``'F'``, ``'X'``.
+"""
+
+import re
+
+import numpy as np
+
+from . import readtext, logic
+from .circuit import Circuit
+
+_pattern_re = re.compile(r'^\s*(?:\d+:\s*)?([01HLRFX-]+)\s*$')
+
+
+class TestFile:
+    """Intermediate representation of a *.test pattern file.
+
+    :param text: The contents of a *.test file.
+    """
+
+    def __init__(self, text: str):
+        patterns = []
+        for line in text.splitlines():
+            if (m := _pattern_re.match(line)) is not None:
+                patterns.append(logic.mvarray(m.group(1)))
+
+        self.patterns = np.stack(patterns, axis=-1) if patterns \
+            else np.zeros((0, 0), dtype=np.uint8)
+        """The parsed patterns as a mvarray (see :py:mod:`~kyupy.logic`).
+
+        The second-to-last axis goes along primary inputs, the last axis goes along patterns.
+        """
+
+    def tests(self, circuit: Circuit):
+        """Assembles and returns the test pattern set for the given circuit.
+
+        :param circuit: The circuit to assemble the patterns for. The patterns will follow the
+            :py:attr:`~kyupy.circuit.Circuit.io_nodes` ordering of the circuit.
+        :return: A logic array (see :py:mod:`~kyupy.logic`). The values for primary inputs are
+            filled, all other values are left unassigned.
+        """
+        pi_locs = [i for i, n in enumerate(circuit.io_nodes) if len(n.ins) == 0]
+        tests = np.full((len(circuit.io_nodes), self.patterns.shape[-1]), logic.UNASSIGNED, dtype=np.uint8)
+        tests[pi_locs] = self.patterns
+        return tests
+
+
+def parse(text) -> TestFile:
+    """Parses the given ``text`` and returns a :class:`TestFile` object."""
+    return TestFile(text)
+
+
+def load(file) -> TestFile:
+    """Parses the contents of ``file`` and returns a :class:`TestFile` object.
+
+    Files with `.gz`-suffix are decompressed on-the-fly.
+    """
+    return parse(readtext(file))

src/kyupy/bench.py

@@ -1,43 +1,118 @@
+"""A parser for the ISCAS89 benchmark format.
+
+The ISCAS89 benchmark format (`.bench`-suffix) is a very simple textual description of gate-level netlists.
+Historically it was first used in the
+`ISCAS89 benchmark set <https://people.engr.ncsu.edu/brglez/CBL/benchmarks/ISCAS89/>`_.
+Besides loading these benchmarks, this module is also useful for easily constructing simple circuits:
+``c = bench.parse('input(x, y) output(a, o, n) a=and(x,y) o=or(x,y) n=not(x)')``.
+
+This parser automatically translates gate to simulation primitives supported by kyupy.
+In most cases this means just adding the number of inputs to the name: e.g. NAND -> NAND2, OR -> OR4, INV -> INV1.
+Since simulators only support at most four-input-gates, gates with more than four inputs are implemented as tree.
+"""
+
 from lark import Lark, Transformer
+
 from .circuit import Circuit, Node, Line
+from . import readtext, batchrange, log
+
+def treeify(l, max_degree=4):
+    if len(l) <= max_degree: return l
+    return treeify([l[bo:bo+bs] for bo, bs in batchrange(len(l), max_degree)])
 
 
 class BenchTransformer(Transformer):
-    
+
     def __init__(self, name):
         super().__init__()
         self.c = Circuit(name)
-    
+        self.inputs = set()
+        self.outputs = set()
+
     def start(self, _): return self.c
-        
-    def parameters(self, args): return [self.c.get_or_add_fork(name) for name in args]
-        
-    def interface(self, args): self.c.interface.extend(args[0])
+
+    def parameters(self, args): return [self.c.get_or_add_fork(str(name)) for name in args if name is not None]
+
+    def input(self, args):
+        for input in args[0]:
+            if input in self.outputs:
+                new_input = self.c.get_or_add_fork(input.name + '~i')
+                Line(self.c, new_input, input)
+                log.warn(f'input-output passthrough, renaming input: {input.name} -> {new_input.name}')
+                input = new_input
+            self.inputs.add(input)
+            self.c.io_nodes.append(input)
+
+    def output(self, args):
+        for output in args[0]:
+            if output in self.inputs:
+                new_output = self.c.get_or_add_fork(output.name + '~o')
+                Line(self.c, output, new_output)
+                log.warn(f'input-output passthrough, renaming output: {output.name} -> {new_output.name}')
+                output = new_output
+            self.outputs.add(output)
+            self.c.io_nodes.append(output)
+
+    def _cell_tree_inner(self, name, kind, inner_kind, drivers):
+        cell = Node(self.c, name, f'{kind}{len(drivers)}')
+        fork = self.c.get_or_add_fork(name)
+        Line(self.c, cell, fork)
+        for i, d in enumerate(drivers):
+            while isinstance(d, list) and len(d) == 1: d = d[0]
+            if isinstance(d, list):
+                d = self._cell_tree_inner(f'{name}~{i}', inner_kind, inner_kind, d)
+            Line(self.c, d, cell)
+        return fork
+
+    def cell_tree(self, name, kind, drivers):
+        root_kind = kind.upper()
+        inner_kind = root_kind
+        if root_kind == 'NAND': inner_kind = 'AND'
+        if root_kind == 'NOR': inner_kind = 'OR'
+        if root_kind == 'XNOR': inner_kind = 'XOR'
+        return self._cell_tree_inner(name, root_kind, inner_kind, treeify(drivers))
 
     def assignment(self, args):
-        name, cell_type, drivers = args
-        cell = Node(self.c, str(name), str(cell_type))
+        name, kind, drivers = args
+        if kind.upper() in ('AND', 'NAND', 'OR', 'NOR', 'XOR', 'XNOR'):
+            self.cell_tree(name, kind, drivers)
+            return
+        if kind.upper().startswith('BUF'):
+            kind = 'BUF1'
+        elif kind.upper().startswith('INV') or kind.upper().startswith('NOT'):
+            kind = 'INV1'
+        cell = Node(self.c, str(name), str(kind))
         Line(self.c, cell, self.c.get_or_add_fork(str(name)))
-        [Line(self.c, d, cell) for d in drivers]
-        
+        for d in drivers: Line(self.c, d, cell)
 
-def parse(bench):
-    grammar = r"""
+
+GRAMMAR = r"""
     start: (statement)*
     statement: input | output | assignment
-    input: ("INPUT" | "input") parameters -> interface
-    output: ("OUTPUT" | "output") parameters -> interface
+    input: ("INPUT" | "input") parameters
+    output: ("OUTPUT" | "output") parameters
     assignment: NAME "=" NAME parameters
     parameters: "(" [ NAME ( "," NAME )* ] ")"
     NAME: /[-_a-z0-9]+/i
     %ignore ( /\r?\n/ | "#" /[^\n]*/ | /[\t\f ]/ )+
     """
-    name = None
-    if '(' not in str(bench):  # No parentheses?: Assuming it is a file name.
-        name = str(bench).replace('.bench', '')
-        with open(bench, 'r') as f:
-            text = f.read()
-    else:
-        text = bench
-    return Lark(grammar, parser="lalr", transformer=BenchTransformer(name)).parse(text)
 
+
+def parse(text, name=None) -> Circuit:
+    """Parses the given ``text`` as ISCAS89 bench code.
+
+    :param text: A string with bench code.
+    :param name: The name of the circuit. Circuit names are not included in bench descriptions.
+    :return: A :class:`Circuit` object.
+    """
+    return Lark(GRAMMAR, parser="lalr", transformer=BenchTransformer(name)).parse(text) # type: ignore
+
+
+def load(file, name=None):
+    """Parses the contents of ``file`` as ISCAS89 bench code.
+
+    :param file: The file to be loaded. Files with `.gz`-suffix are decompressed on-the-fly.
+    :param name: The name of the circuit. If None, the file name is used as circuit name.
+    :return: A :class:`Circuit` object.
+    """
+    return parse(readtext(file), name=name or str(file))

src/kyupy/bittools.py

@@ -1,23 +0,0 @@
-import numpy as np
-import importlib.util
-if importlib.util.find_spec('numba') is not None:
-    import numba
-else:
-    from . import numba
-    print('Numba unavailable. Falling back to pure python')
-
-
-_pop_count_lut = np.asarray([bin(x).count('1') for x in range(256)])
-
-
-def popcount(a):
-    return np.sum(_pop_count_lut[a])
-
-
-_bit_in_lut = np.array([2 ** x for x in range(7, -1, -1)], dtype='uint8')
-
-
-@numba.njit
-def bit_in(a, pos):
-    return a[pos >> 3] & _bit_in_lut[pos & 7]
-

src/kyupy/circuit.py

@@ -1,12 +1,54 @@
-from collections import deque
+"""Core module for handling non-hierarchical gate-level circuits.
 
+The class :class:`Circuit` is a container of nodes connected by lines.
+A node is an instance of class :class:`Node`,
+and a line is an instance of class :class:`Line`.
 
-class GrowingList(list):
+The data structures are designed to work together nicely with numpy arrays.
+For example, all the nodes and connections in the circuit graph have consecutive integer indices that can be used to access ndarrays with associated data.
+Circuit graphs also define an ordering of inputs, outputs and other nodes to easily process test vector data and alike.
+
+"""
+
+from __future__ import annotations
+
+from collections import deque, defaultdict
+import re
+
+import numpy as np
+
+from . import NestedNumericList, NestedStrIntDict
+
+class GrowingList[T](list[T]):
     def __setitem__(self, index, value):
-        if index >= len(self):
-            self.extend([None] * (index + 1 - len(self)))
+        if value is None: self.has_nones = True
+        if index == len(self): return super().append(value)
+        if index > len(self):
+            super().extend([None] * (index + 1 - len(self)))
+            self.has_nones = True
         super().__setitem__(index, value)
 
+    # Override __getitem__ to return None when reading beyond the list
+    # instead of throwing an exception. Type checker complains about the None return
+    # type, though. Probably not needed anyways.
+
+    #def __getitem__(self, index) -> list[T] | T | None:
+    #    if isinstance(index, slice): return super().__getitem__(index)
+    #    return super().__getitem__(index) if index < len(self) else None
+
+    @property
+    def free_idx(self):
+        fi = len(self)
+        if hasattr(self, 'has_nones') and self.has_nones:
+            fi = next((i for i, x in enumerate(self) if x is None), len(self))
+            self.has_nones = fi < len(self)
+        return fi
+
+    def without_nones(self):
+        for item in self:
+            if item is not None:
+                yield item
+
 
 class IndexList(list):
     def __delitem__(self, index):
@@ -19,43 +61,70 @@ class IndexList(list):
 
 
 class Node:
-    """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.
-    Each node contains:
+    """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 is connected to other nodes via lines.
 
-    * `self.index`: a circuit-unique integer index.
-    * `self.kind`: a type describing its function (e.g. 'AND', 'NOR').
-      The type '__fork__' is special. It signifies a named signal
-      or a fan-out in the circuit. Any other type is considered a physical cell.
-    * `self.name`: a name. Names must be unique among all forks and all cells
-      in the circuit. However, a fork (`self.kind=='__fork__'`) and a cell with
-      the same name may coexist.
-    * `self.ins`: a list of input connections (objects of class `Line`)
-    * `self.outs`: a list of output connections (objects of class `Line`).
+    The constructor automatically adds the new node to the given circuit.
     """
     def __init__(self, circuit, name, kind='__fork__'):
         if kind == '__fork__':
-            if name in circuit.forks:
-                raise ValueError(f'fork of name {name} already exists.')
+            assert name not in circuit.forks, f'fork of name {name} already in circuit.'
             circuit.forks[name] = self
         else:
-            if name in circuit.cells:
-                raise ValueError(f'cell of name {name} already exists.')
+            assert name not in circuit.cells, f'cell of name {name} already in circuit.'
             circuit.cells[name] = self
-        self.index = len(circuit.nodes)
         circuit.nodes.append(self)
         self.circuit = circuit
+        """The :class:`Circuit` object the node is part of.
+        """
         self.name = name
+        """The name of the node.
+
+        Names must be unique among all forks and all cells in the circuit.
+        However, a fork (:py:attr:`kind` is set to '__fork__') and a cell with the same name may coexist.
+        """
         self.kind = kind
-        self.ins = GrowingList()
-        self.outs = GrowingList()
+        """A string describing the type of the node.
+
+        Common types are the names from a standard cell library or general gate names like 'AND' or 'NOR'.
+        If :py:attr:`kind` is set to '__fork__', it receives special treatment.
+        A `fork` describes a named signal or a fan-out point in the circuit and not a physical `cell` like a gate.
+        In the circuit, the namespaces of forks and cells are kept separate.
+        While :py:attr:`name` must be unique among all forks and all cells, a fork can have the same name as a cell.
+        The :py:attr:`index`, however, is unique among all nodes; a fork cannot have the same index as a cell.
+        """
+        self.index = len(circuit.nodes) - 1
+        """A unique and consecutive integer index of the node within the circuit.
+
+        It can be used to associate additional data to a node :code:`n`
+        by allocating an array or list :code:`my_data` of length :code:`len(n.circuit.nodes)` and
+        accessing it by :code:`my_data[n.index]` or simply by :code:`my_data[n]`.
+        """
+        self.ins: GrowingList[Line] = GrowingList()
+        """A list of input connections (:class:`Line` objects).
+        """
+        self.outs: GrowingList[Line] = GrowingList()
+        """A list of output connections (:class:`Line` objects).
+        """
+
+    def __index__(self):
+        return self.index
 
     def __repr__(self):
         ins = ' '.join([f'<{line.index}' if line is not None else '<None' for line in self.ins])
         outs = ' '.join([f'>{line.index}' if line is not None else '>None' for line in self.outs])
-        return f'{self.index}:{self.kind}"{self.name}" {ins} {outs}'
+        ins = ' ' + ins if len(ins) else ''
+        outs = ' ' + outs if len(outs) else ''
+        return f'{self.index}:{self.kind}"{self.name}"{ins}{outs}'
 
     def remove(self):
+        """Removes the node from its circuit.
+
+        Lines may still reference the removed node.
+        The user must connect such lines to other nodes or remove the lines from the circuit.
+        To keep the indices consecutive, the node with the highest index within the circuit
+        will be assigned the index of the removed node.
+        """
         if self.circuit is not None:
             del self.circuit.nodes[self.index]
             if self.kind == '__fork__':
@@ -64,58 +133,84 @@ class Node:
                 del self.circuit.cells[self.name]
             self.circuit = None
 
+    def __eq__(self, other):
+        """Checks equality of node name and kind. Does not check pin connections.
+
+        This is ok, because (name, kind) is unique within a circuit.
+        """
+        return self.name == other.name and self.kind == other.kind and self.circuit == other.circuit
+
+    def __hash__(self):
+        return hash((self.name, self.kind, id(self.circuit)))
+
 
 class Line:
-    """A Line is a directional 1:1 connection between two Nodes. It always
-    connects an output of a node (called `driver`) to an input of a node
-    (called `reader`) and has a circuit-unique index (`self.index`).
+    """A line is a directional 1:1 connection between two nodes.
 
-    Furthermore, `self.driver_pin` and `self.reader_pin` are the
-    integer indices of the connected pins of the nodes. They always correspond
-    to the positions of the line in the connection lists of the nodes:
+    It always connects an output of one `driver` node to an input of one `reader` node.
+    If a signal fans out to multiple readers, a '__fork__' node needs to be added.
 
-    * `self.driver.outs[self.driver_pin] == self`
-    * `self.reader.ins[self.reader_pin] == self`
+    The constructor automatically adds the new line to the given circuit and inserts references into the connection
+    lists of connected nodes.
 
-    A Line always connects a single driver to a single reader. If a signal fans out to
-    multiple readers, a '__fork__' Node needs to be added.
+    When adding a line, input and output pins can either be specified explicitly
+    :code:`Line(circuit, (driver, 2), (reader, 0))`, or implicitly :code:`Line(circuit, driver, reader)`.
+    In the implicit case, the line will be connected to the first free pin of the node.
+    Use the explicit case only if connections to specific pins are required.
+    It may overwrite any previous line references in the connection list of the nodes.
     """
-    def __init__(self, circuit, driver, reader):
-        self.index = len(circuit.lines)
-        circuit.lines.append(self)
-        if type(driver) is Node:
-            self.driver = driver
-            self.driver_pin = len(driver.outs)
-            for pin, line in enumerate(driver.outs):
-                if line is None:
-                    self.driver_pin = pin
-                    break
-        else:
-            self.driver, self.driver_pin = driver
-        if type(reader) is Node:
-            self.reader = reader
-            self.reader_pin = len(reader.ins)
-            for pin, line in enumerate(reader.ins):
-                if line is None:
-                    self.reader_pin = pin
-                    break
-        else:
-            self.reader, self.reader_pin = reader
+    def __init__(self, circuit: Circuit, driver: Node | tuple[Node, None|int], reader: Node | tuple[Node, None|int]):
+        self.circuit = circuit
+        """The :class:`Circuit` object the line is part of.
+        """
+        self.circuit.lines.append(self)
+        self.index = len(self.circuit.lines) - 1
+        """A unique and consecutive integer index of the line within the circuit.
+
+        It can be used to store additional data about the line :code:`l`
+        by allocating an array or list :code:`my_data` of length :code:`len(l.circuit.lines)` and
+        accessing it by :code:`my_data[l.index]` or simply by :code:`my_data[l]`.
+        """
+        if not isinstance(driver, tuple): driver = (driver, driver.outs.free_idx)
+        self.driver: Node = driver[0]
+        """The :class:`Node` object that drives this line.
+        """
+        self.driver_pin = driver[1] if driver[1] is not None else self.driver.outs.free_idx
+        """The output pin position of the driver node this line is connected to.
+
+        This is the position in the list :py:attr:`Node.outs` of the driving node this line referenced from:
+        :code:`self.driver.outs[self.driver_pin] == self`.
+        """
+        if not isinstance(reader, tuple): reader = (reader, reader.ins.free_idx)
+        self.reader: Node = reader[0]
+        """The :class:`Node` object that reads this line.
+        """
+        self.reader_pin = reader[1] if reader[1] is not None else self.reader.ins.free_idx
+        """The input pin position of the reader node this line is connected to.
+
+        This is the position in the list :py:attr:`Node.ins` of the reader node this line referenced from:
+        :code:`self.reader.ins[self.reader_pin] == self`.
+        """
         self.driver.outs[self.driver_pin] = self
         self.reader.ins[self.reader_pin] = self
 
     def remove(self):
-        circuit = None
+        """Removes the line from its circuit and its referencing nodes.
+
+        To keep the indices consecutive, the line with the highest index within the circuit
+        will be assigned the index of the removed line.
+        """
         if self.driver is not None:
             self.driver.outs[self.driver_pin] = None
-            circuit = self.driver.circuit
-        if self.reader is not None:
-            self.reader.ins[self.reader_pin] = None
-            circuit = self.reader.circuit
-        if circuit is not None:
-            del circuit.lines[self.index]
-        self.driver = None
-        self.reader = None
+            if self.driver.kind == '__fork__':  # squeeze outputs
+                del self.driver.outs[self.driver_pin]
+                for i, l in enumerate(self.driver.outs): l.driver_pin = i
+        if self.reader is not None: self.reader.ins[self.reader_pin] = None
+        if self.circuit is not None: del self.circuit.lines[self.index]
+        self.circuit = None
+
+    def __index__(self):
+        return self.index
 
     def __repr__(self):
         return f'{self.index}'
@@ -123,31 +218,290 @@ class Line:
     def __lt__(self, other):
         return self.index < other.index
 
+    def __eq__(self, other):
+        return self.driver == other.driver and self.driver_pin == other.driver_pin and \
+               self.reader == other.reader and self.reader_pin == other.reader_pin
+
+    def __hash__(self):
+        return hash((self.driver, self.driver_pin, self.reader, self.reader_pin))
+
 
 class Circuit:
     """A Circuit is a container for interconnected nodes and lines.
 
-    All contained lines have unique indices, so have all contained nodes.
-    These indices can be used to store additional data about nodes or lines
-    by allocating an array `my_data` of length `len(self.nodes)` and then
-    accessing it by `my_data[n.index]`. The indices may change iff lines or
-    nodes are removed from the circuit.
+    It provides access to lines by index and to nodes by index and by name.
+    Nodes come in two flavors: `cells` and `forks` (see :py:attr:`Node.kind`).
+    The name spaces of cells and forks are kept separate.
 
-    Nodes come in two flavors (cells and forks, see `Node`). The names of
-    these nodes are kept unique within these two flavors.
+    The indices of nodes and lines are kept consecutive and unique.
+    Whenever lines or nodes are removed from the circuit, the indices of some other lines or nodes may change
+    to enforce consecutiveness.
+
+    A subset of nodes can be designated as primary input- or output-ports of the circuit.
+    This is done by adding them to the :py:attr:`io_nodes` list.
     """
     def __init__(self, name=None):
         self.name = name
-        self.nodes = IndexList()
-        self.lines = IndexList()
-        self.interface = GrowingList()
-        self.cells = {}
-        self.forks = {}
+        """The name of the circuit.
+        """
+        self.nodes : list[Node] = IndexList()
+        """A list of all :class:`Node` objects contained in the circuit.
+
+        The position of a node in this list equals its index :code:`self.nodes[42].index == 42`.
+        This list must not be changed directly.
+        Use the :class:`Node` constructor and :py:attr:`Node.remove()` to add and remove nodes.
+        """
+        self.lines : list[Line] = IndexList()
+        """A list of all :class:`Line` objects contained in the circuit.
+
+        The position of a line in this list equals its index :code:`self.lines[42].index == 42`.
+        This list must not be changed directly.
+        Use the :class:`Line` constructor and :py:attr:`Line.remove()` to add and remove lines.
+        """
+        self.io_nodes : list[Node] = GrowingList()
+        """A list of nodes that are designated as primary input- or output-ports.
+
+        Port-nodes are contained in :py:attr:`nodes` as well as :py:attr:`io_nodes`.
+        The position of a node in the io_nodes list corresponds to positions of logic values in test vectors.
+        The port direction is not stored explicitly.
+        Usually, nodes in the io_nodes list without any lines in their :py:attr:`Node.ins` list are primary inputs,
+        and all other nodes in the io_nodes list are regarded as primary outputs.
+        """
+        self.cells : dict[str, Node] = {}
+        """A dictionary to access cells by name.
+
+        This dictionary must not be changed directly.
+        Use the :class:`Node` constructor and :py:attr:`Node.remove()` to add and remove nodes.
+        """
+        self.forks : dict[str, Node] = {}
+        """A dictionary to access forks by name.
+
+        This dictionary must not be changed directly.
+        Use the :class:`Node` constructor and :py:attr:`Node.remove()` to add and remove nodes.
+        """
+
+    def s_nodes(self, tlib: 'TechLib') -> list: # type: ignore
+        """A list of all primary I/Os as well as all flip-flops and latches in the circuit (in that order).
+
+        The s_nodes list defines the order of all ports and all sequential elements in the circuit.
+        This list is constructed on-the-fly. If used in some inner toop, consider caching the list for better performance.
+        """
+        return list(self.io_nodes) + [n for n in self.nodes if tlib.is_dff(n.kind)] + [n for n in self.nodes if tlib.is_latch(n.kind)]
+
+    def io_locs(self, prefix):
+        """Returns a list of indices of primary I/Os that start with given name prefix.
+
+        The returned values are used to index into the :py:attr:`io_nodes` array.
+        If only one I/O cell matches the given prefix, a list with a single integer is returned.
+        If a bus matches the given prefix, a sorted list of indices is returned.
+        Busses are identified by integers in the cell names following the given prefix.
+        Lists for bus indices are sorted from LSB (e.g. :code:`data[0]`) to MSB (e.g. :code:`data[31]`).
+        If a prefix matches multiple different signals or busses, alphanumerically sorted
+        lists of lists are returned. Therefore, higher-dimensional busses
+        (e.g. :code:`data0[0], data0[1], ...`, :code:`data1[0], data1[1], ...`) are supported as well.
+        """
+        return self._locs(prefix, list(self.io_nodes))
+
+    def s_locs(self, prefix, tlib: 'TechLib'): # type: ignore
+        """Returns the indices of I/Os and sequential elements that start with given name prefix.
+
+        The returned values are used to index into the :py:attr:`s_nodes` list.
+        It works the same as :py:attr:`io_locs`. See there for more details.
+        """
+        return self._locs(prefix, self.s_nodes(tlib))
+
+    def _locs(self, prefix, nodes:list[Node]) -> NestedNumericList:  # can return list[list[...]]
+        d_top: NestedStrIntDict = dict()
+        for i, n in enumerate(nodes):
+            if m := re.match(fr'({re.escape(prefix)}.*?)((?:[\d_\[\]])*$)', n.name):
+                path = [m[1]] + [int(v) for v in re.split(r'[_\[\]]+', m[2]) if len(v) > 0]
+                d = d_top
+                for j in path[:-1]:
+                    d[j] = d.get(j, dict())
+                    d = d[j]
+                    assert isinstance(d, dict)
+                d[path[-1]] = i
+
+        # sort recursively for multi-dimensional lists.
+        def sorted_values(d) -> NestedNumericList: return [sorted_values(v) for k, v in sorted(d.items())] if isinstance(d, dict) else d
+        l = sorted_values(d_top)
+        while isinstance(l, list) and len(l) == 1 and isinstance(l[0], list): l = l[0]
+        return l
+
+    def stats(self, tlib: 'TechLib'): # type: ignore
+        """A dictionary with the counts of all different elements in the circuit.
+
+        The dictionary contains the number of all different kinds of nodes, the number
+        of lines, as well various sums like number of combinational gates, number of
+        primary I/Os, number of sequential elements, and so on.
+
+        The count of regular cells use their :py:attr:`Node.kind` as key, other statistics use
+        dunder-keys like: `__comb__`, `__io__`, `__seq__`, and so on.
+        """
+        stats = defaultdict(int)
+        stats['__node__'] = len(self.nodes)
+        stats['__cell__'] = len(self.cells)
+        stats['__fork__'] = len(self.forks)
+        stats['__io__'] = len(self.io_nodes)
+        stats['__line__'] = len(self.lines)
+        for n in self.cells.values():
+            stats[n.kind] += 1
+            if tlib.is_dff(n.kind): stats['__dff__'] += 1
+            elif tlib.is_latch(n.kind): stats['__latch__'] += 1
+            elif 'put' not in n.kind.lower(): stats['__comb__'] += 1 # don't count input or output towards comb
+        stats['__seq__'] = stats['__dff__'] + stats['__latch__']
+        return dict(stats)
 
     def get_or_add_fork(self, name):
         return self.forks[name] if name in self.forks else Node(self, name)
-    
+
+    def remove_dangling_nodes(self, root_node:Node, keep=[]):
+        if len([l for l in root_node.outs if l is not None]) > 0: return
+        lines = [l for l in root_node.ins if l is not None]
+        drivers = [l.driver for l in lines]
+        if root_node in keep: return
+        root_node.remove()
+        for l in lines:
+            l.remove()
+        for d in drivers:
+            self.remove_dangling_nodes(d, keep=keep)
+
+    def eliminate_1to1_forks(self):
+        """Removes all forks that drive only one node.
+
+        Such forks are inserted by parsers to annotate signal names. If this
+        information is not needed, such forks can be removed and the two neighbors
+        can be connected directly using one line. Forks that drive more than one node
+        are not removed by this function.
+
+        This function may remove some nodes and some lines from the circuit.
+        Therefore that indices of other nodes and lines may change to keep the indices consecutive.
+        It may therefore invalidate external data for nodes and lines.
+        """
+        ios = set(self.io_nodes)
+        for n in list(self.forks.values()):
+            if n in ios: continue
+            if len(n.outs) != 1: continue
+            in_line = n.ins[0]
+            out_line = n.outs[0]
+            out_reader = out_line.reader
+            out_reader_pin = out_line.reader_pin
+            n.remove()
+            out_line.remove()
+            in_line.reader = out_reader
+            in_line.reader_pin = out_reader_pin
+            in_line.reader.ins[in_line.reader_pin] = in_line
+
+    def remove_forks(self):
+        ios = set(self.io_nodes)
+        for n in list(self.forks.values()):
+            if n in ios: continue
+            d = None
+            if (l := n.ins[0]) is not None:
+                d = l.driver
+                l.remove()
+            for l in list(n.outs):
+                if l is None: continue
+                r, rp = l.reader, l.reader_pin
+                l.remove()
+                if d is not None: Line(self, d, (r, rp))
+            n.remove()
+
+    def substitute(self, node, impl):
+        """Replaces a given node with the given implementation circuit.
+
+        The given node will be removed, the implementation is copied in and
+        the signal lines are connected appropriately. The number and arrangement
+        of the input and output ports must match the pins of the replaced node.
+
+        This function tries to preserve node and line indices as much as possible.
+        Usually, it only adds additional nodes and lines, preserving the order of
+        all existing nodes and lines. If an implementation is empty, however, nodes
+        and lines may get removed, changing indices and invalidating external data.
+        """
+        ios = set(impl.io_nodes)
+        impl_in_nodes = [n for n in impl.io_nodes if len(n.ins) == 0]
+        impl_out_lines = [n.ins[0] for n in impl.io_nodes if len(n.ins) > 0]
+        designated_cell = None
+        if len(impl_out_lines) > 0:
+            n = impl_out_lines[0].driver
+            while n.kind == '__fork__' and n not in ios:
+                n = n.ins[0].driver
+            designated_cell = n
+        node_in_lines = list(node.ins) + [None] * (len(impl_in_nodes)-len(node.ins))
+        node_out_lines = list(node.outs) + [None] * (len(impl_out_lines)-len(node.outs))
+        assert len(node_in_lines) == len(impl_in_nodes)
+        assert len(node_out_lines) == len(impl_out_lines)
+        node_map = dict()
+        if designated_cell is not None:
+            node.kind = designated_cell.kind
+            node_map[designated_cell] = node
+            node.ins = GrowingList()
+            node.outs = GrowingList()
+        else:
+            node.remove()
+        ios = set(impl.io_nodes)
+        for n in impl.nodes:  # add all nodes to main circuit
+            if n not in ios:
+                if n != designated_cell:
+                    node_map[n] = Node(self, f'{node.name}~{n.name}', n.kind)
+            elif len(n.outs) > 0 and len(n.ins) > 0:  # output is also read by impl. circuit, need to add a fork.
+                node_map[n] = Node(self, f'{node.name}~{n.name}')
+            elif len(n.ins) == 0 and len(n.outs) != 1:  # input is read by multiple nodes (or no nodes), need to add fork.
+                node_map[n] = Node(self, f'{node.name}~{n.name}')
+        for l in impl.lines:  # add all internal lines to main circuit
+            if l.reader in node_map and l.driver in node_map:
+                Line(self, (node_map[l.driver], l.driver_pin), (node_map[l.reader], l.reader_pin))
+        for inn, ll in zip(impl_in_nodes, node_in_lines):  # connect inputs
+            if ll is None: continue
+            if len(inn.outs) == 1:
+                l = inn.outs[0]
+                ll.reader = node_map[l.reader]
+                ll.reader_pin = l.reader_pin
+            else:
+                ll.reader = node_map[inn]  # connect to existing fork
+                ll.reader_pin = 0
+            ll.reader.ins[ll.reader_pin] = ll
+        for l, ll in zip(impl_out_lines, node_out_lines):  # connect outputs
+            if ll is None:
+                if l.driver in node_map:
+                    self.remove_dangling_nodes(node_map[l.driver], keep=ios)
+                continue
+            if len(l.reader.outs) > 0:  # output is also read by impl. circuit, connect to fork.
+                ll.driver = node_map[l.reader]
+                ll.driver_pin = len(l.reader.outs)
+            else:
+                ll.driver = node_map[l.driver]
+                ll.driver_pin = l.driver_pin
+            ll.driver.outs[ll.driver_pin] = ll
+
+    def resolve_tlib_cells(self, tlib):
+        """Substitute all technology library cells with kyupy native simulation primitives.
+
+        See :py:attr:`substitute()` for more detail.
+        """
+        for n in list(self.nodes):
+            if n.kind in tlib.cells:
+                self.substitute(n, tlib.cells[n.kind][0])
+
+    def remove_constants(self):
+        c1gen = None
+        for n in self.nodes:
+            if n.kind == '__const0__':  # just remove, unconnected inputs are defined 0.
+                for l in n.outs:
+                    l.remove()
+                n.remove()
+            elif n.kind == '__const1__':
+                if c1gen is None: c1gen = Node(self, '__const1gen__', 'INV1')  # one unique const 1 generator
+                for l in n.outs:
+                    r, rp = l.reader, l.reader_pin
+                    l.remove()
+                    Line(self, c1gen, (r, rp))
+                n.remove()
+
     def copy(self):
+        """Returns a deep copy of the circuit.
+        """
         c = Circuit(self.name)
         for node in self.nodes:
             Node(c, node.name, node.kind)
@@ -155,82 +509,181 @@ class Circuit:
             d = c.forks[line.driver.name] if line.driver.kind == '__fork__' else c.cells[line.driver.name]
             r = c.forks[line.reader.name] if line.reader.kind == '__fork__' else c.cells[line.reader.name]
             Line(c, (d, line.driver_pin), (r, line.reader_pin))
-        for node in self.interface:
+        for node in self.io_nodes:
             if node.kind == '__fork__':
                 n = c.forks[node.name]
             else:
                 n = c.cells[node.name]
-            c.interface.append(n)
+            c.io_nodes.append(n)
         return c
-    
-    def dump(self):
-        header = f'{self.name}({",".join([str(n.index) for n in self.interface])})\n'
-        return header + '\n'.join([str(n) for n in self.nodes])
+
+    def __getstate__(self):
+        nodes = [(node.name, node.kind) for node in self.nodes]
+        lines = [(line.driver.index, line.driver_pin, line.reader.index, line.reader_pin) for line in self.lines]
+        io_nodes = [n.index for n in self.io_nodes]
+        return {'name': self.name,
+                'nodes': nodes,
+                'lines': lines,
+                'io_nodes': io_nodes }
+
+    def __setstate__(self, state):
+        self.name = state['name']
+        self.nodes = IndexList()
+        self.lines = IndexList()
+        self.io_nodes = GrowingList()
+        self.cells = {}
+        self.forks = {}
+        for s in state['nodes']:
+            Node(self, *s)
+        for driver, driver_pin, reader, reader_pin in state['lines']:
+            Line(self, (self.nodes[driver], driver_pin), (self.nodes[reader], reader_pin))
+        for n in state['io_nodes']:
+            self.io_nodes.append(self.nodes[n])
 
     def __repr__(self):
-        name = f" '{self.name}'" if self.name else ''
-        return f'<Circuit{name} with {len(self.nodes)} nodes, {len(self.lines)} lines, {len(self.interface)} ports>'
+        return f'{{name: "{self.name}", cells: {len(self.cells)}, forks: {len(self.forks)}, lines: {len(self.lines)}, io_nodes: {len(self.io_nodes)}}}'
 
     def topological_order(self):
-        visit_count = [0] * len(self.nodes)
-        queue = deque(n for n in self.nodes if len(n.ins) == 0 or 'DFF' in n.kind)
+        """Generator function to iterate over all nodes in topological order.
+
+        Nodes without input lines and nodes whose :py:attr:`Node.kind` contains the
+        substrings 'dff' or 'latch' are yielded first.
+        """
+        visit_count = np.zeros(len(self.nodes), dtype=np.uint32)
+        start = set(n for n in self.nodes if len(n.ins) == 0 or 'dff' in n.kind.lower() or 'latch' in n.kind.lower())
+        queue = deque(start)
         while len(queue) > 0:
             n = queue.popleft()
             for line in n.outs:
                 if line is None: continue
                 succ = line.reader
-                visit_count[succ.index] += 1
-                if visit_count[succ.index] == len(succ.ins) and 'DFF' not in succ.kind:
+                visit_count[succ] += 1
+                if visit_count[succ] == len(succ.ins) and succ not in start:
                     queue.append(succ)
             yield n
 
+    def topological_order_with_level(self):
+        level = np.zeros(len(self.nodes), dtype=np.int32) - 1
+        for n in self.topological_order():
+            if len(n.ins) == 0 or 'dff' in n.kind.lower() or 'latch' in n.kind.lower():
+                l = 0
+            else:
+                l = level[[l.driver.index for l in n.ins if l is not None]].max() + 1
+            level[n] = l
+            yield n, l
+
     def topological_line_order(self):
+        """Generator function to iterate over all lines in topological order.
+        """
         for n in self.topological_order():
             for line in n.outs:
                 if line is not None:
                     yield line
 
-    def reversed_topological_order(self):
+    def reversed_topological_order(self, tlib: 'TechLib'): # type: ignore
+        """Generator function to iterate over all nodes in reversed topological order.
+
+        Nodes without output lines and sequential nodes (flip-flops, latches) are yielded first.
+        """
         visit_count = [0] * len(self.nodes)
-        queue = deque(n for n in self.nodes if len(n.outs) == 0 or 'DFF' in n.kind)
+        queue = deque(n for n in self.nodes if len(n.outs) == 0 or tlib.is_dff(n.kind) or tlib.is_latch(n.kind))
         while len(queue) > 0:
             n = queue.popleft()
             for line in n.ins:
                 pred = line.driver
-                visit_count[pred.index] += 1
-                if visit_count[pred.index] == len(pred.outs) and 'DFF' not in pred.kind:
+                visit_count[pred] += 1
+                if visit_count[pred] == len(pred.outs) and not tlib.is_dff(pred.kind) and not tlib.is_latch(pred.kind):
                     queue.append(pred)
             yield n
 
-    def fanin(self, origin_nodes):
+    def fanin(self, origin_nodes, tlib: 'TechLib'): # type: ignore
+        """Generator function to iterate over the fan-in cone of a given list of origin nodes.
+
+        Nodes are yielded in reversed topological order.
+        """
         marks = [False] * len(self.nodes)
         for n in origin_nodes:
-            marks[n.index] = True
-        for n in self.reversed_topological_order():
-            if not marks[n.index]:
-                for line in n.outs:
-                    if line is not None:
-                        marks[n.index] |= marks[line.reader.index]
-            if marks[n.index]:
+            marks[n] = True
+        for n in self.reversed_topological_order(tlib):
+            if not marks[n]:
+                for line in n.outs.without_nones():
+                    marks[n] |= marks[line.reader]
+            if marks[n]:
                 yield n
 
-    def fanout_free_regions(self):
-        for stem in self.reversed_topological_order():
-            if len(stem.outs) == 1 and 'DFF' not in stem.kind: continue
+    def fanout(self, origin_nodes: list[Node], node_filter = lambda n: 'dff' not in n.name.lower()):
+        """Generator function to iterate over the fan-out cone of a given list of origin nodes.
+
+        origin_nodes are yielded first, followed by nodes driven by them in a breadth-first manner.
+        The search stops at nodes for which node_filter returns False.
+        Only origin_nodes and nodes for which node_filter returned True are yielded.
+        By default, search stops at flip-flops.
+        """
+        queue = deque(origin_nodes)
+        yielded = set()
+        while len(queue) > 0:
+            n = queue.popleft()
+            for line in n.outs.without_nones():
+                succ = line.reader
+                if succ not in yielded and node_filter(succ):
+                    yielded.add(succ)
+                    queue.append(succ)
+            yield n
+
+    def fanout_free_regions(self, tlib: 'TechLib'): # type: ignore
+        for stem in self.reversed_topological_order(tlib):
+            if len(stem.outs) == 1 and 'dff' not in stem.kind.lower(): continue
             region = []
-            if 'DFF' in stem.kind:
+            if 'dff' in stem.kind.lower():
                 n = stem.ins[0]
-                if len(n.driver.outs) == 1 and 'DFF' not in n.driver.kind:
+                if len(n.driver.outs) == 1 and 'dff' not in n.driver.kind.lower():
                     queue = deque([n.driver])
                 else:
                     queue = deque()
             else:
                 queue = deque(n.driver for n in stem.ins
-                              if len(n.driver.outs) == 1 and 'DFF' not in n.driver.kind)
+                              if len(n.driver.outs) == 1 and 'dff' not in n.driver.kind.lower())
             while len(queue) > 0:
                 n = queue.popleft()
                 preds = [pred.driver for pred in n.ins
-                         if len(pred.driver.outs) == 1 and 'DFF' not in pred.driver.kind]
+                         if len(pred.driver.outs) == 1 and 'dff' not in pred.driver.kind.lower()]
                 queue.extend(preds)
                 region.append(n)
             yield stem, region
+
+    def dot(self, tlib: 'TechLib', format='svg', graph_attr={}, line_labels={}):
+        from graphviz import Digraph
+        dot = Digraph(format=format, graph_attr={'rankdir': 'LR', 'splines': 'true', 'size': '10', 'ranksep': '0.1'} | graph_attr)
+
+        s_dict = dict((n, i) for i, n in enumerate(self.s_nodes(tlib)))
+        node_level = np.zeros(len(self.nodes), dtype=np.uint32)
+        level_nodes = defaultdict(list)
+        for n, lv in self.topological_order_with_level():
+            level_nodes[lv].append(n)
+            node_level[n] = lv
+
+        for lv in level_nodes:
+            with dot.subgraph() as s:
+                s.attr(rank='same')
+                for n in level_nodes[lv]:
+                    ins = '{' + '|'.join([f'<i{i}>{i}' for i in range(len(n.ins))]) + '}|' if len(n.ins) > 1 else ''
+                    outs = '|{' + '|'.join([f'<o{i}>{i}' for i in range(len(n.outs))]) + '}' if len(n.outs) > 1 else ''
+                    io = f' [{s_dict[n]}]' if n in s_dict else ''
+                    color = '#f5f5f5' if n.kind == '__fork__' else '#cccccc'
+                    kind = '' if n.kind == '__fork__' else fr'\n{n.kind}'
+                    s.node(name=str(n.index), label = fr'{{{ins}{n.index}{io}{kind}\n{n.name}{outs}}}', shape='record', style='filled', fillcolor=color)
+
+        for l in self.lines:
+            driver = f'{l.driver.index}:o{l.driver_pin}' if len(l.driver.outs)>1 else f'{l.driver.index}'
+            reader = f'{l.reader.index}:i{l.reader_pin}' if len(l.reader.ins)>1 else f'{l.reader.index}'
+            label = str(line_labels.get(l, l.index))
+            if node_level[l.driver] == node_level[l.reader]:
+                dot.node(f'_{l.index}_')
+                dot.edge(driver, f'_{l.index}_', style='dotted', label=label)
+                dot.edge(f'_{l.index}_', reader, style='dotted', label=label)
+            elif node_level[l.driver] > node_level[l.reader]:
+                dot.edge(driver, reader, style='dotted', label=label)
+            else:
+                dot.edge(driver, reader, label=label)
+
+        return dot

src/kyupy/def_file.py

@@ -0,0 +1,297 @@
+"""A simple and incomplete parser for the Design Exchange Format (DEF).
+
+This parser extracts information on components and nets from DEF files and make them available
+as an intermediate representation (:class:`DefFile` object).
+"""
+
+from collections import defaultdict
+
+from lark import Lark, Transformer, Tree
+
+from kyupy import readtext
+
+
+class DefNet:
+    def __init__(self, name):
+        self.name = name
+        self.pins = []
+
+    @property
+    def wires(self):
+        ww = defaultdict(list)
+        [ww[dw.layer].append((int(dw.width), dw.wire_points)) for dw in self.routed if len(dw.wire_points) > 0]
+        return ww
+
+    @property
+    def vias(self):
+        vv = defaultdict(list)
+        [vv[vtype].extend(locs) for dw in self.routed for vtype, locs in dw.vias.items()]
+        return vv
+
+
+class DefWire:
+    def __init__(self):
+        self.layer = None
+        self.width = None
+        self.points = []
+
+    @property
+    def wire_points(self):
+        start = [self.points[0]]
+        rest = [p for p in self.points[1:] if not isinstance(p[0], str)]  # skip over vias
+        return start + rest if len(rest) > 0 else []
+
+    @property
+    def vias(self):
+        vv = defaultdict(list)
+        loc = self.points[0]
+        for p in self.points[1:]:
+            if not isinstance(p[0], str):  # new location
+                loc = (loc[0] if p[0] is None else p[0], loc[1] if p[1] is None else p[1])  # if None, keep previous value
+                continue
+            vtype, param = p
+            if isinstance(param, tuple):  # expand "DO x BY y STEP xs ys"
+                x_cnt, y_cnt, x_sp, y_sp = param
+                [vv[vtype].append((loc[0] + x*x_sp, loc[1] + y*y_sp, 'N')) for x in range(x_cnt) for y in range(y_cnt)]
+            else:
+                vv[vtype].append((loc[0], loc[1], param or 'N'))
+        return vv
+
+    def __repr__(self):
+        return f'<DefWire {self.layer} {self.width} {self.points}>'
+
+
+class DefVia:
+    def __init__(self, name):
+        self.name = name
+        self.rowcol = [1, 1]
+        self.cutspacing = [0, 0]
+
+
+class DefPin:
+    def __init__(self, name):
+        self.name = name
+        self.points = []
+
+
+class DefFile:
+    """Intermediate representation of a DEF file."""
+    def __init__(self):
+        self.rows = []
+        self.tracks = []
+        self.units = []
+        self.vias = {}
+        self.components = {}
+        self.pins = {}
+        self.specialnets = {}
+        self.nets = {}
+
+
+class DefTransformer(Transformer):
+    def __init__(self): self.def_file = DefFile()
+    def start(self, args): return self.def_file
+    def design(self, args): self.def_file.design = args[0].value
+    def point(self, args): return tuple(int(arg.value) if arg != '*' else None for arg in args)
+    def do_step(self, args): return tuple(map(int, args))
+    def spnet_wires(self, args): return args[0].lower(), args[1:]
+    def net_wires(self, args): return args[0].lower(), args[1:]
+    def sppoints(self, args): return args
+    def points(self, args): return args
+    def net_pin(self, args): return '__pin__', (args[0].value, args[1].value)
+    def net_opt(self, args): return args[0].lower(), args[1].value
+
+    def file_stmt(self, args):
+        value = args[1].value
+        value = value[1:-1] if value[0] == '"' else value
+        setattr(self.def_file, args[0].lower(), value)
+
+    def design_stmt(self, args):
+        stmt = args[0].lower()
+        if stmt == 'units': self.def_file.units.append((args[1].value, args[2].value, int(args[3])))
+        elif stmt == 'diearea': self.def_file.diearea = args[1:]
+        elif stmt == 'row':
+            self.def_file.rows.append((args[1].value,  # rowName
+                                       args[2].value,  # siteName
+                                       (int(args[3]), int(args[4])),  # origin x/y
+                                       args[5].value,  # orientation
+                                       max(args[6][0], args[6][1]),  # number of sites
+                                       max(args[6][2], args[6][3])  # site width
+                                      ))
+        elif stmt == 'tracks':
+            self.def_file.tracks.append((args[1].value,  # orientation
+                                         int(args[2]),  # start
+                                         int(args[3]),  # number of tracks
+                                         int(args[4]),  # spacing
+                                         args[5].value  # layer
+                                        ))
+
+    def vias_stmt(self, args):
+        via = DefVia(args[0].value)
+        [setattr(via, opt, val) for opt, val in args[1:]]
+        self.def_file.vias[via.name] = via
+
+    def vias_opt(self, args):
+        opt = args[0].lower()
+        if opt in ['viarule', 'pattern']: val = args[1].value
+        elif opt in ['layers']: val = [arg.value for arg in args[1:]]
+        else: val = [int(arg) for arg in args[1:]]
+        return opt, val
+
+    def comp_stmt(self, args):
+        name = args[0].value
+        kind = args[1].value
+        point = args[2]
+        orientation = args[3].value
+        self.def_file.components[name] = (kind, point, orientation)
+
+    def pins_stmt(self, args):
+        pin = DefPin(args[0].value)
+        [pin.points.append(val) if opt == 'placed' else setattr(pin, opt, val) for opt, val in args[1:]]
+        self.def_file.pins[pin.name] = pin
+
+    def pins_opt(self, args):
+        opt = args[0].lower()
+        if opt in ['net', 'direction', 'use']: val = args[1].value
+        elif opt in ['layer']: val = [args[1].value] + args[2:]
+        elif opt in ['placed']: val = (args[1][0], args[1][1], args[2].value)
+        else: val = []
+        return opt, val
+
+    def spnets_stmt(self, args):
+        dnet = DefNet(args[0].value)
+        for arg in args[1:]:
+            if arg[0] == '__pin__': dnet.pins.append(arg[1])
+            else: setattr(dnet, arg[0], arg[1])
+        self.def_file.specialnets[dnet.name] = dnet
+
+    def nets_stmt(self, args):
+        dnet = DefNet(args[0].value)
+        for arg in args[1:]:
+            if arg[0] == '__pin__': dnet.pins.append(arg[1])
+            else: setattr(dnet, arg[0], arg[1])
+        self.def_file.nets[dnet.name] = dnet
+
+    def spwire(self, args):
+        wire = DefWire()
+        wire.layer = args[0].value
+        wire.width = args[1].value
+        wire.points = args[-1]
+        return wire
+
+    def wire(self, args):
+        wire = DefWire()
+        wire.layer = args[0].value
+        wire.points = args[-1]
+        return wire
+
+    def sppoints_via(self, args):
+        if len(args) == 1: return args[0].value, None
+        else: return args[0].value, args[1]
+
+    def points_via(self, args):
+        if len(args) == 1: return args[0].value, 'N'
+        else: return args[0].value, args[1].value.strip()
+
+
+GRAMMAR = r"""
+    start: /#[^\n]*/? file_stmt*
+
+    ?file_stmt: /VERSION/ ID ";"
+              | /DIVIDERCHAR/ STRING ";"
+              | /BUSBITCHARS/ STRING ";"
+              | design
+
+    design: "DESIGN" ID ";" design_stmt* "END" "DESIGN"
+
+    ?design_stmt: /UNITS/ ID ID NUMBER ";"
+                | /DIEAREA/ point+ ";"
+                | /ROW/ ID ID NUMBER NUMBER ID do_step ";"
+                | /TRACKS/ /[XY]/ NUMBER "DO" NUMBER "STEP" NUMBER "LAYER" ID ";"
+                | propdef | vias | nondef | comp | pins | pinprop | spnets | nets
+
+    propdef: "PROPERTYDEFINITIONS" propdef_stmt* "END" "PROPERTYDEFINITIONS"
+    propdef_stmt: /COMPONENTPIN/ ID ID ";"
+
+    vias: "VIAS" NUMBER ";" vias_stmt* "END" "VIAS"
+    vias_stmt: "-" ID vias_opt* ";"
+    vias_opt: "+" /VIARULE/ ID
+            | "+" /CUTSIZE/ NUMBER NUMBER
+            | "+" /LAYERS/ ID ID ID
+            | "+" /CUTSPACING/ NUMBER NUMBER
+            | "+" /ENCLOSURE/ NUMBER NUMBER NUMBER NUMBER
+            | "+" /ROWCOL/ NUMBER NUMBER
+            | "+" /PATTERN/ ID
+
+    nondef: "NONDEFAULTRULES" NUMBER ";" nondef_stmt+ "END" "NONDEFAULTRULES"
+    nondef_stmt: "-" ID ( "+" /HARDSPACING/
+                        | "+" /LAYER/ ID "WIDTH" NUMBER "SPACING" NUMBER
+                        | "+" /VIA/ ID )* ";"
+
+    comp: "COMPONENTS" NUMBER ";" comp_stmt* "END" "COMPONENTS"
+    comp_stmt: "-" ID ID "+" "PLACED" point ID ";"
+
+    pins: "PINS" NUMBER ";" pins_stmt* "END" "PINS"
+    pins_stmt: "-" ID pins_opt* ";"
+    pins_opt: "+" /NET/ ID
+            | "+" /SPECIAL/
+            | "+" /DIRECTION/ ID
+            | "+" /USE/ ID
+            | "+" /PORT/
+            | "+" /LAYER/ ID point point
+            | "+" /PLACED/ point ID
+
+    pinprop: "PINPROPERTIES" NUMBER ";" pinprop_stmt* "END" "PINPROPERTIES"
+    pinprop_stmt: "-" "PIN" ID "+" "PROPERTY" ID STRING ";"
+
+    spnets: "SPECIALNETS" NUMBER ";" spnets_stmt* "END" "SPECIALNETS"
+    spnets_stmt: "-" ID ( net_pin | net_opt | spnet_wires )* ";"
+
+    spnet_wires: "+" ( /COVER/ | /FIXED/ | /ROUTED/ ) spwire ( "NEW" spwire )*
+
+    spwire: ID NUMBER spwire_opt* sppoints
+    spwire_opt: "+" /SHAPE/ ID
+              | "+" /STYLE/ ID
+
+    sppoints: point ( point | sppoints_via )+
+    sppoints_via: ID do_step?
+
+    nets: "NETS" NUMBER ";" nets_stmt* "END" "NETS"
+    nets_stmt: "-" ID ( net_pin | net_opt | net_wires )* ";"
+
+    net_pin: "(" ID ID ")"
+    net_opt: "+" /USE/ ID
+           | "+" /NONDEFAULTRULE/ ID
+    net_wires: "+" ( /COVER/ | /FIXED/ | /ROUTED/ | /NOSHIELD/ ) wire ( "NEW" wire )*
+
+    wire: ID wire_opt points
+    wire_opt: ( "TAPER" | "TAPERRULE" ID )? ("STYLE" ID)?
+
+    points: point ( point | points_via )+
+    points_via: ID ORIENTATION?
+
+    point: "(" (NUMBER|/\*/) (NUMBER|/\*/) NUMBER? ")"
+
+    do_step: "DO" NUMBER "BY" NUMBER "STEP" (NUMBER|SIGNED_NUMBER) (NUMBER|SIGNED_NUMBER)
+
+    ORIENTATION.2: /F?[NWES]/ WS
+    ID: /[^ \t\f\r\n+][^ \t\f\r\n]*/
+    STRING : "\"" /.*?/s /(?<!\\)(\\\\)*?/ "\""
+    WS: /[ \t\f\r\n]/
+
+    %import common.NUMBER
+    %import common.SIGNED_NUMBER
+    %ignore WS (/#[^\n]*/)?
+    """
+
+
+def parse(text):
+    """Parses the given ``text`` and returns a :class:`DefFile` object."""
+    return Lark(GRAMMAR, parser="lalr", transformer=DefTransformer()).parse(text)
+
+
+def load(file):
+    """Parses the contents of ``file`` and returns a :class:`DefFile` object.
+
+    Files with `.gz`-suffix are decompressed on-the-fly.
+    """
+    return parse(readtext(file))

src/kyupy/logic.py

@@ -0,0 +1,481 @@
+"""Core module for handling 2-, 4-, and 8-valued logic data and signal values.
+
+Logic values are stored in numpy arrays with data type ``np.uint8``.
+There are no explicit data structures in KyuPy for holding patterns, pattern sets or vectors.
+However, there are conventions on logic value encoding and on the order of axes.
+Utility functions defined here follow these conventions.
+
+8 logic values are defined as integer constants.
+
+* For 2-valued logic: ``ZERO`` and ``ONE``
+* 4-valued logic adds: ``UNASSIGNED`` and ``UNKNOWN``
+* 8-valued logic adds: ``RISE``, ``FALL``, ``PPULSE``, and ``NPULSE``.
+
+In general, the bits in these constants have the following meaning:
+
+* bit0: Final/settled binary value of a signal
+* bit1: Initial binary value of a signal
+* bit2: Activity or transitions are present on a signal
+
+Except when bit0 differs from bit1, but bit2 (activity) is 0:
+
+* bit0 = 1, bit1 = 0, bit2 = 0 means ``UNKNOWN`` in 4-valued and 8-valued logic.
+* bit0 = 0, bit1 = 1, bit2 = 0 means ``UNASSIGNED`` in 4-valued and 8-valued logic.
+
+2-valued logic only considers bit0, but should store logic one as ``ONE=0b011`` for interoperability.
+4-valued logic only considers bit0 and bit1.
+8-valued logic considers all 3 bits.
+
+Logic values are stored in numpy arrays of data type ``np.uint8``.
+The axis convention is as follows:
+
+* The **last** axis goes along patterns/vectors. I.e. ``values[...,0]`` is pattern 0, ``values[...,1]`` is pattern 1, etc.
+* The **second-to-last** axis goes along the I/O and flip-flops of circuits. For a circuit ``c``, this axis is usually
+  ``len(c.s_nodes)`` long. The values of all inputs, outputs and flip-flops are stored within the same array and the location
+  along the second-to-last axis is determined by the order in :py:attr:`~kyupy.circuit.Circuit.s_nodes`.
+
+Two storage formats are used in KyuPy:
+
+* ``mv...`` (for "multi-valued"): Each logic value is stored in the least significant 3 bits of ``np.uint8``.
+* ``bp...`` (for "bit-parallel"): Groups of 8 logic values are stored as three ``np.uint8``. This format is used
+  for bit-parallel logic simulations. It is also more memory-efficient.
+
+The functions in this module use the ``mv...`` and ``bp...`` prefixes to signify the storage format they operate on.
+
+"""
+
+from collections.abc import Iterable
+
+import numpy as np
+from numpy.typing import DTypeLike
+
+from . import numba, hr_bytes, NestedNumericList
+
+
+ZERO = 0b000
+"""Integer constant ``0b000`` for logic-0. ``'0'``, ``0``, ``False``, ``'L'``, and ``'l'`` are interpreted as ``ZERO``.
+"""
+UNKNOWN = 0b001
+"""Integer constant ``0b001`` for unknown or conflict. ``'X'``, or any other value is interpreted as ``UNKNOWN``.
+"""
+UNASSIGNED = 0b010
+"""Integer constant ``0b010`` for unassigned or high-impedance. ``'-'``, ``None``, ``'Z'``, and ``'z'`` are
+interpreted as ``UNASSIGNED``.
+"""
+ONE = 0b011
+"""Integer constant ``0b011`` for logic-1. ``'1'``, ``1``, ``True``, ``'H'``, and ``'h'`` are interpreted as ``ONE``.
+"""
+PPULSE = 0b100
+"""Integer constant ``0b100`` for positive pulse, meaning initial and final values are 0, but there is some activity
+on a signal. ``'P'``, ``'p'``, and ``'^'`` are interpreted as ``PPULSE``.
+"""
+RISE = 0b101
+"""Integer constant ``0b110`` for a rising transition. ``'R'``, ``'r'``, and ``'/'`` are interpreted as ``RISE``.
+"""
+FALL = 0b110
+"""Integer constant ``0b101`` for a falling transition. ``'F'``, ``'f'``, and ``'\\'`` are interpreted as ``FALL``.
+"""
+NPULSE = 0b111
+"""Integer constant ``0b111`` for negative pulse, meaning initial and final values are 1, but there is some activity
+on a signal. ``'N'``, ``'n'``, and ``'v'`` are interpreted as ``NPULSE``.
+"""
+
+
+#                      0 X - 1 P R F N
+_mv_xor   = np.array([[0,1,1,3,4,5,6,7], # 0
+                      [1,1,1,1,1,1,1,1], # X
+                      [1,1,1,1,1,1,1,1], # -
+                      [3,1,1,0,7,6,5,4], # 1
+                      [4,1,1,7,4,5,6,7], # P
+                      [5,1,1,6,5,4,7,6], # R
+                      [6,1,1,5,6,7,4,5], # F
+                      [7,1,1,4,7,6,5,4], # N
+                      ], dtype=np.uint8)
+
+#                      0 X - 1 P R F N
+_mv_merge = np.array([[0,0,0,1,1,1,1,1], # 0
+                      [0,1,1,3,4,5,6,7], # X
+                      [0,1,2,3,4,5,6,7], # -
+                      [1,3,3,3,1,1,1,1], # 1
+                      [1,4,4,1,4,1,1,1], # P
+                      [1,5,5,1,1,5,1,1], # R
+                      [1,6,6,1,1,1,6,1], # F
+                      [1,7,7,1,1,1,1,7], # N
+                      ], dtype=np.uint8)
+
+
+def interpret(value) -> int|NestedNumericList:
+    """Converts characters, strings, and lists of them to lists of logic constants defined above.
+
+    :param value: A character (string of length 1), Boolean, Integer, None, or Iterable.
+        Iterables (such as strings) are traversed and their individual characters are interpreted.
+    :return: A logic constant or a (possibly multi-dimensional) list of logic constants.
+    """
+    if isinstance(value, Iterable) and not (isinstance(value, str) and len(value) == 1):
+        return list(map(interpret, value))
+    if value in [0, '0', False, 'L', 'l']: return ZERO
+    if value in [1, '1', True, 'H', 'h']: return ONE
+    if value in [None, '-', 'Z', 'z']: return UNASSIGNED
+    if value in ['R', 'r', '/']: return RISE
+    if value in ['F', 'f', '\\']: return FALL
+    if value in ['P', 'p', '^']: return PPULSE
+    if value in ['N', 'n', 'v']: return NPULSE
+    return UNKNOWN
+
+
+def mvarray(*a) -> np.ndarray[tuple[int]|tuple[int,int],np.dtype[np.uint8]]:
+    """Converts (lists of) Boolean values or strings into a multi-valued array.
+
+    Iterable over characters, Booleans, Integers (0,1), None are interpreted as a single pattern.
+    Multi-character strings are Iterables over characters and are interpreted as a single pattern.
+    Lists of said patterns form a pattern set as ndarray with the axes are arranged as per KyuPy's convention.
+    """
+    mva = np.array(interpret(a), dtype=np.uint8)
+    if mva.ndim < 2: return mva
+    if mva.shape[-2] > 1: return mva.swapaxes(-1, -2)
+    return mva[..., 0, :]
+
+
+def mv_str(mva, delim='\n'):
+    """Renders a given multi-valued array into a string.
+    """
+    sa = np.choose(mva, np.array([*'0X-1PRFN'], dtype=np.str_))
+    if not hasattr(mva, 'ndim') or mva.ndim == 0: return sa
+    if mva.ndim == 1: return ''.join(sa)
+    return delim.join([''.join(c) for c in sa.swapaxes(-1,-2)])
+
+
+def _mv_not(out, inp):
+    np.bitwise_xor(inp, 0b11, out=out)  # this also exchanges UNASSIGNED <-> UNKNOWN
+    np.putmask(out, (inp == UNKNOWN), UNKNOWN)  # restore UNKNOWN
+
+
+def mv_not(x1 : np.ndarray, out=None):
+    """A multi-valued NOT operator.
+
+    :param x1: A multi-valued array.
+    :param out: An optional storage destination. If None, a new multi-valued array is returned.
+    :return: A multi-valued array with the result.
+    """
+    out = out or np.empty(x1.shape, dtype=np.uint8)
+    _mv_not(out, x1)
+    return out
+
+
+def _mv_or(out, *ins):
+    any_unknown = (ins[0] == UNKNOWN) | (ins[0] == UNASSIGNED)
+    for inp in ins[1:]: any_unknown |= (inp == UNKNOWN) | (inp == UNASSIGNED)
+    any_one = (ins[0] == ONE)
+    for inp in ins[1:]: any_one |= (inp == ONE)
+
+    out[...] = ZERO
+    np.putmask(out, any_one, ONE)
+    for inp in ins:
+        np.bitwise_or(out, inp, out=out, where=~any_one)
+    np.putmask(out, (any_unknown & ~any_one), UNKNOWN)
+
+
+def mv_or(x1, x2, out=None):
+    """A multi-valued OR operator.
+
+    :param x1: A multi-valued array.
+    :param x2: A multi-valued array.
+    :param out: An optional storage destination. If None, a new multi-valued array is returned.
+    :return: A multi-valued array with the result.
+    """
+    out = out or np.empty(np.broadcast(x1, x2).shape, dtype=np.uint8)
+    _mv_or(out, x1, x2)
+    return out
+
+
+def _mv_and(out, *ins):
+    any_unknown = (ins[0] == UNKNOWN) | (ins[0] == UNASSIGNED)
+    for inp in ins[1:]: any_unknown |= (inp == UNKNOWN) | (inp == UNASSIGNED)
+    any_zero = (ins[0] == ZERO)
+    for inp in ins[1:]: any_zero |= (inp == ZERO)
+
+    out[...] = ONE
+    np.putmask(out, any_zero, ZERO)
+    for inp in ins:
+        np.bitwise_and(out, inp | 0b100, out=out, where=~any_zero)
+        np.bitwise_or(out, inp & 0b100, out=out, where=~any_zero)
+    np.putmask(out, (any_unknown & ~any_zero), UNKNOWN)
+
+
+def mv_and(x1, x2, out=None):
+    """A multi-valued AND operator.
+
+    :param x1: A multi-valued array.
+    :param x2: A multi-valued array.
+    :param out: An optional storage destination. If None, a new multi-valued array is returned.
+    :return: A multi-valued array with the result.
+    """
+    out = out or np.empty(np.broadcast(x1, x2).shape, dtype=np.uint8)
+    _mv_and(out, x1, x2)
+    return out
+
+
+def mv_xor(x1, x2, out=None):
+    """A multi-valued XOR operator.
+
+    :param x1: A multi-valued array.
+    :param x2: A multi-valued array.
+    :param out: An optional storage destination. If None, a new multi-valued array is returned.
+    :return: A multi-valued array with the result.
+    """
+    out = out or np.empty(np.broadcast(x1, x2).shape, dtype=np.uint8)
+    out[...] = _mv_xor[x1, x2]
+    return out
+
+
+def mv_merge(x1, x2, out=None):
+    """A multi-valued merge operator.
+
+    Merging identical values ``a`` and ``a`` yields ``a``.
+    Merging ``X`` and ``-`` yields ``X``.
+    Merging any value ``a`` with ``X`` or ``-`` yields ``a``.
+    Merging any other values ``a != b`` yields ``X``.
+
+    :param x1: A multi-valued array.
+    :param x2: A multi-valued array.
+    :param out: An optional storage destination. If None, a new multi-valued array is returned.
+    :return: A multi-valued array with the result.
+    """
+    out = out or np.empty(np.broadcast(x1, x2).shape, dtype=np.uint8)
+    out[...] = _mv_merge[x1, x2]
+    return out
+
+
+def mv_latch(d, t, q_prev, out=None):
+    """A multi-valued latch operator.
+
+    A latch outputs ``d`` when transparent (``t`` is high).
+    It outputs ``q_prev`` when in latched state (``t`` is low).
+
+    :param d: A multi-valued array for the data input.
+    :param t: A multi-valued array for the control input.
+    :param q_prev: A multi-valued array with the output value of this latch from the previous clock cycle.
+    :param out: An optional storage destination. If None, a new multi-valued array is returned.
+    :return: A multi-valued array for the latch output ``q``.
+    """
+    out = out or np.empty(np.broadcast(d, t, q_prev).shape, dtype=np.uint8)
+    out[...] = t & d & 0b011
+    out[...] |= ~t & 0b010 & (q_prev << 1)
+    out[...] |= ~t & 0b001 & (out >> 1)
+    out[...] |= ((out << 1) ^ (out << 2)) & 0b100
+    unknown = (t == UNKNOWN) \
+              | (t == UNASSIGNED) \
+              | (((d == UNKNOWN) | (d == UNASSIGNED)) & (t != ZERO))
+    np.putmask(out, unknown, UNKNOWN)
+    return out
+
+
+def mv_transition(init, final, out=None):
+    """Computes the logic transitions from the initial values of ``init`` to the final values of ``final``.
+    Pulses in the input data are ignored. If any of the inputs are ``UNKNOWN``, the result is ``UNKNOWN``.
+    If init is ``UNASSIGNED``, the result is the final value of ``final``.
+    If final is ``UNASSIGNED``, the result is the initial value of ``init``.
+    If both inputs are ``UNASSIGNED``, the result is ``UNASSIGNED``.
+
+    :param init: A multi-valued array.
+    :param final: A multi-valued array.
+    :param out: An optional storage destination. If None, a new multi-valued array is returned.
+    :return: A multi-valued array with the result.
+    """
+    out = out or np.empty(np.broadcast(init, final).shape, dtype=np.uint8)
+    out[...] = (init & 0b010) | (final & 0b001)
+    out[...] |= ((out << 1) ^ (out << 2)) & 0b100
+    out[...] = np.choose(init == UNASSIGNED, [out, (final & 0b001) * ONE])
+    out[...] = np.choose(final == UNASSIGNED, [out, ((init & 0b010) >> 1) * ONE])
+    unknown = (init == UNKNOWN) | (final == UNKNOWN)
+    unassigned = (init == UNASSIGNED) & (final == UNASSIGNED)
+    np.putmask(out, unknown, UNKNOWN)
+    np.putmask(out, unassigned, UNASSIGNED)
+    return out
+
+
+def mv_to_bp(mva):
+    """Converts a multi-valued array into a bit-parallel array.
+    """
+    if mva.ndim == 1: mva = mva[..., np.newaxis]
+    return np.packbits(unpackbits(mva)[...,:3], axis=-2, bitorder='little').swapaxes(-1,-2)
+
+
+def mv_init(mva):
+    """Returns the initial binary values for mva.
+    """
+    return (mva>>1) & ((mva>>2)|mva) & 1
+
+
+def mv_final(mva):
+    """Returns the final binary value of mva.
+    """
+    return mva & ((mva>>2)|(mva>>1)) & 1
+
+
+def bparray(*a):
+    """Converts (lists of) Boolean values or strings into a bit-parallel array.
+
+    The given values are interpreted and the axes are arranged as per KyuPy's convention.
+    Use this function to convert strings into bit-parallel arrays.
+    """
+    return mv_to_bp(mvarray(*a))
+
+
+def bp_to_mv(bpa):
+    """Converts a bit-parallel array into a multi-valued array.
+    """
+    return packbits(np.unpackbits(bpa, axis=-1, bitorder='little').swapaxes(-1,-2))
+
+
+def bp4v_buf(out, inp):
+    unknown = inp[..., 0, :] ^ inp[..., 1, :]
+    out[..., 0, :] = inp[..., 0, :] | unknown
+    out[..., 1, :] = inp[..., 1, :] & ~unknown
+    return out
+
+
+def bp8v_buf(out, inp):
+    unknown = (inp[..., 0, :] ^ inp[..., 1, :]) & ~inp[..., 2, :]
+    out[..., 0, :] = inp[..., 0, :] | unknown
+    out[..., 1, :] = inp[..., 1, :] & ~unknown
+    out[..., 2, :] = inp[..., 2, :] & ~unknown
+    return out
+
+
+def bp4v_not(out, inp):
+    unknown = inp[..., 0, :] ^ inp[..., 1, :]
+    out[..., 0, :] = ~inp[..., 0, :] | unknown
+    out[..., 1, :] = ~inp[..., 1, :] & ~unknown
+    return out
+
+
+def bp8v_not(out, inp):
+    unknown = (inp[..., 0, :] ^ inp[..., 1, :]) & ~inp[..., 2, :]
+    out[..., 0, :] = ~inp[..., 0, :] | unknown
+    out[..., 1, :] = ~inp[..., 1, :] & ~unknown
+    out[..., 2, :] = inp[..., 2, :] & ~unknown
+    return out
+
+
+def bp4v_or(out, *ins):
+    out[...] = 0
+    any_unknown = ins[0][..., 0, :] ^ ins[0][..., 1, :]
+    for inp in ins[1:]: any_unknown |= inp[..., 0, :] ^ inp[..., 1, :]
+    any_one = ins[0][..., 0, :] & ins[0][..., 1, :]
+    for inp in ins[1:]: any_one |= inp[..., 0, :] & inp[..., 1, :]
+    for inp in ins:
+        out[..., 0, :] |= inp[..., 0, :] | any_unknown
+        out[..., 1, :] |= inp[..., 1, :] & (~any_unknown | any_one)
+    return out
+
+
+def bp8v_or(out, *ins):
+    out[...] = 0
+    any_unknown = (ins[0][..., 0, :] ^ ins[0][..., 1, :]) & ~ins[0][..., 2, :]
+    for inp in ins[1:]: any_unknown |= (inp[..., 0, :] ^ inp[..., 1, :]) & ~inp[..., 2, :]
+    any_one = ins[0][..., 0, :] & ins[0][..., 1, :] & ~ins[0][..., 2, :]
+    for inp in ins[1:]: any_one |= inp[..., 0, :] & inp[..., 1, :] & ~inp[..., 2, :]
+    for inp in ins:
+        out[..., 0, :] |= inp[..., 0, :] | any_unknown
+        out[..., 1, :] |= inp[..., 1, :] & (~any_unknown | any_one)
+        out[..., 2, :] |= inp[..., 2, :] & (~any_unknown | any_one) & ~any_one
+    return out
+
+
+def bp4v_and(out, *ins):
+    out[...] = 0xff
+    any_unknown = ins[0][..., 0, :] ^ ins[0][..., 1, :]
+    for inp in ins[1:]: any_unknown |= inp[..., 0, :] ^ inp[..., 1, :]
+    any_zero = ~ins[0][..., 0, :] & ~ins[0][..., 1, :]
+    for inp in ins[1:]: any_zero |= ~inp[..., 0, :] & ~inp[..., 1, :]
+    for inp in ins:
+        out[..., 0, :] &= inp[..., 0, :] | (any_unknown & ~any_zero)
+        out[..., 1, :] &= inp[..., 1, :] & ~any_unknown
+    return out
+
+
+def bp8v_and(out, *ins):
+    out[...] = 0xff
+    any_unknown = (ins[0][..., 0, :] ^ ins[0][..., 1, :]) & ~ins[0][..., 2, :]
+    for inp in ins[1:]: any_unknown |= (inp[..., 0, :] ^ inp[..., 1, :]) & ~inp[..., 2, :]
+    any_zero = ~ins[0][..., 0, :] & ~ins[0][..., 1, :] & ~ins[0][..., 2, :]
+    for inp in ins[1:]: any_zero |= ~inp[..., 0, :] & ~inp[..., 1, :] & ~inp[..., 2, :]
+    out[..., 2, :] = 0
+    for inp in ins:
+        out[..., 0, :] &= inp[..., 0, :] | (any_unknown & ~any_zero)
+        out[..., 1, :] &= inp[..., 1, :] & ~any_unknown
+        out[..., 2, :] |= inp[..., 2, :] & (~any_unknown | any_zero) & ~any_zero
+    return out
+
+
+def bp4v_xor(out, *ins):
+    out[...] = 0
+    any_unknown = ins[0][..., 0, :] ^ ins[0][..., 1, :]
+    for inp in ins[1:]: any_unknown |= inp[..., 0, :] ^ inp[..., 1, :]
+    for inp in ins:
+        out[..., 0, :] ^= inp[..., 0, :]
+        out[..., 1, :] ^= inp[..., 1, :]
+    out[..., 0, :] |= any_unknown
+    out[..., 1, :] &= ~any_unknown
+    return out
+
+
+def bp8v_xor(out, *ins):
+    out[...] = 0
+    any_unknown = (ins[0][..., 0, :] ^ ins[0][..., 1, :]) & ~ins[0][..., 2, :]
+    for inp in ins[1:]: any_unknown |= (inp[..., 0, :] ^ inp[..., 1, :]) & ~inp[..., 2, :]
+    for inp in ins:
+        out[..., 0, :] ^= inp[..., 0, :]
+        out[..., 1, :] ^= inp[..., 1, :]
+        out[..., 2, :] |= inp[..., 2, :]
+    out[..., 0, :] |= any_unknown
+    out[..., 1, :] &= ~any_unknown
+    out[..., 2, :] &= ~any_unknown
+    return out
+
+
+def bp8v_latch(out, d, t, q_prev):
+    any_unknown = (t[..., 0, :] ^ t[..., 1, :]) & ~t[..., 2, :]
+    any_unknown |= ((d[..., 0, :] ^ d[..., 1, :]) & ~d[..., 2, :]) & (t[..., 0, :] | t[..., 1, :] | t[..., 2, :])
+    out[..., 1, :] = (d[..., 1, :] & t[..., 1, :]) | (q_prev[..., 0, :] & ~t[..., 1, :])
+    out[..., 0, :] = (d[..., 0, :] & t[..., 0, :]) | (out[..., 1, :] & ~t[..., 0, :])
+    out[..., 2, :] = out[..., 1, :] ^ out[..., 0, :]
+    out[..., 0, :] |= any_unknown
+    out[..., 1, :] &= ~any_unknown
+    out[..., 2, :] &= ~any_unknown
+    return out
+
+
+_bit_in_lut = np.array([2 ** x for x in range(7, -1, -1)], dtype='uint8')
+
+
+@numba.njit
+def bit_in(a, pos):
+    return a[pos >> 3] & _bit_in_lut[pos & 7]
+
+
+def unpackbits(a : np.ndarray):
+    """Unpacks the bits of given ndarray ``a``.
+
+    Similar to ``np.unpackbits``, but accepts any dtype, preserves the shape of ``a`` and
+    adds a new last axis with the bits of each item. Bits are in 'little'-order, i.e.,
+    a[...,0] is the least significant bit of each item.
+    """
+    return np.unpackbits(a.view(np.uint8), bitorder='little').reshape(*a.shape, 8*a.itemsize)
+
+
+def packbits(a, dtype:DTypeLike=np.uint8):
+    """Packs the values of a boolean-valued array ``a`` along its last axis into bits.
+
+    Similar to ``np.packbits``, but returns an array of given dtype and the shape of ``a`` with the last axis removed.
+    The last axis of `a` is truncated or padded according to the bit-width of the given dtype.
+    Signed integer datatypes are padded with the most significant bit, all others are padded with `0`.
+    """
+    dtype = np.dtype(dtype)
+    bits = 8 * dtype.itemsize
+    a = a[...,:bits]
+    if a.shape[-1] < bits:
+        p = [(0,0)]*(len(a.shape)-1) + [(0, bits-a.shape[-1])]
+        a = np.pad(a, p, 'edge') if dtype.name[0] == 'i' else np.pad(a, p, 'constant', constant_values=0)
+    return np.packbits(a, bitorder='little').view(dtype).reshape(a.shape[:-1])

src/kyupy/packed_vectors.py

@@ -1,299 +0,0 @@
-import numpy as np
-from .bittools import popcount, bit_in
-
-
-class PackedVectors:
-    def __init__(self, nvectors=8, width=1, vdim=1, from_cache=None):
-        if from_cache is not None:
-            self.bits = np.array(from_cache)
-            self.width, self.vdim, nbytes = self.bits.shape
-        else:
-            self.bits = np.zeros((width, vdim, (nvectors - 1) // 8 + 1), dtype='uint8')
-            self.vdim = vdim
-            self.width = width
-        self.nvectors = nvectors
-        m1 = np.array([2 ** x for x in range(7, -1, -1)], dtype='uint8')
-        m0 = ~m1
-        self.mask = np.rollaxis(np.vstack((m0, m1)), 1)
-
-    @classmethod
-    def from_pair(cls, init, final):
-        assert init.nvectors == final.nvectors
-        assert len(init.bits) == len(final.bits)
-        init_v = init.bits[:, 0]
-        if init.vdim == 3:
-            init_c = (init.bits[:, 0] ^ init.bits[:, 1]) | init.bits[:, 2]
-        elif init.vdim == 2:
-            init_c = init.bits[:, 1]
-        else:
-            init_c = ~np.zeros_like(init.bits[:, 0])
-        final_v = final.bits[:, 0]
-        if final.vdim == 3:
-            final_c = (final.bits[:, 0] ^ final.bits[:, 1]) | final.bits[:, 2]
-            final_v = ~final.bits[:, 1]
-        elif final.vdim == 2:
-            final_c = final.bits[:, 1]
-        else:
-            final_c = ~np.zeros_like(final.bits[:, 0])
-        c = init_c & final_c
-        a0 = init_v & c
-        a1 = ~final_v & c
-        a2 = (init_v ^ final_v) & c
-        p = PackedVectors(init.nvectors, len(init.bits), 3)
-        p.bits[:, 0] = a0
-        p.bits[:, 1] = a1
-        p.bits[:, 2] = a2
-        return p
-        
-    def transition_vectors(self):
-        a = PackedVectors(self.nvectors-1, self.width, 3)
-        for pos in range(self.width):
-            for vidx in range(self.nvectors-1):
-                tr = self.get_value(vidx, pos) + self.get_value(vidx+1, pos)
-                if tr == '00':
-                    a.set_value(vidx, pos, '0')
-                elif tr == '11':
-                    a.set_value(vidx, pos, '1')
-                elif tr == '01':
-                    a.set_value(vidx, pos, 'R')
-                elif tr == '10':
-                    a.set_value(vidx, pos, 'F')
-                elif tr == '--':
-                    a.set_value(vidx, pos, '-')
-                else:
-                    a.set_value(vidx, pos, 'X')
-        return a
-        
-    def __add__(self, other):
-        a = PackedVectors(self.nvectors + other.nvectors, self.width, max(self.vdim, other.vdim))
-        # a.bits[:self.bits.shape[0], 0] = self.bits[:, 0]
-        # if self.vdim == 2:
-        #    a.bits[:self.bits.shape[0], 1] = self.care_bits
-        # elif self.vdim == 3:
-        #    a.bits[:self.bits.shape[0], 1] = ~self.value_bits
-        #    a.bits[:self.bits.shape[0], 2] = self.toggle_bits
-        for i in range(self.nvectors):
-            a[i] = self[i]
-        for i in range(len(other)):
-            a[self.nvectors+i] = other[i]
-        return a
-
-    def __len__(self):
-        return self.nvectors
-    
-    def randomize(self, one_probability=0.5):
-        for data in self.bits:
-            data[0] = np.packbits((np.random.rand(self.nvectors) < one_probability).astype(int))
-            if self.vdim == 2:
-                data[1] = 255
-            elif self.vdim == 3:
-                data[1] = ~np.packbits((np.random.rand(self.nvectors) < one_probability).astype(int))
-                data[2] = data[0] ^ ~data[1]
-            
-    def copy(self, selection_mask=None):
-        if selection_mask is not None:
-            cpy = PackedVectors(popcount(selection_mask), len(self.bits), self.vdim)
-            cur = 0
-            for vidx in range(self.nvectors):
-                if bit_in(selection_mask, vidx):
-                    cpy[cur] = self[vidx]
-                    cur += 1
-        else:
-            cpy = PackedVectors(self.nvectors, len(self.bits), self.vdim)
-            np.copyto(cpy.bits, self.bits)
-        return cpy
-
-    @property
-    def care_bits(self):
-        if self.vdim == 1:
-            return self.bits[:, 0] | 255
-        elif self.vdim == 2:
-            return self.bits[:, 1]
-        elif self.vdim == 3:
-            return (self.bits[:, 0] ^ self.bits[:, 1]) | self.bits[:, 2]
-
-    @property
-    def initial_bits(self):
-        return self.bits[:, 0]
-
-    @property
-    def value_bits(self):
-        if self.vdim == 3:
-            return ~self.bits[:, 1]
-        else:
-            return self.bits[:, 0]
-
-    @property
-    def toggle_bits(self):
-        if self.vdim == 3:
-            return self.bits[:, 2]
-        else:
-            return self.bits[:, 0] & 0
-
-    def get_value(self, vector, position):
-        if vector >= self.nvectors:
-            raise IndexError(f'vector out of range: {vector} >= {self.nvectors}')
-        a = self.bits[position, :, vector // 8]
-        m = self.mask[vector % 8]
-        if self.vdim == 1:
-            return '1' if a[0] & m[1] else '0'
-        elif self.vdim == 2:
-            if a[0] & m[1]:
-                return '1' if a[1] & m[1] else 'X'
-            else:
-                return '0' if a[1] & m[1] else '-'
-        elif self.vdim == 3:
-            if a[2] & m[1]:
-                if a[0] & m[1]:
-                    return 'F' if a[1] & m[1] else 'N'
-                else:
-                    return 'P' if a[1] & m[1] else 'R'
-            else:
-                if a[0] & m[1]:
-                    return 'X' if a[1] & m[1] else '1'
-                else:
-                    return '0' if a[1] & m[1] else '-'                
-
-    def get_values_for_position(self, position):
-        return ''.join(self.get_value(x, position) for x in range(self.nvectors))
-
-    def set_value(self, vector, position, v):
-        if vector >= self.nvectors:
-            raise IndexError(f'vector out of range: {vector} >= {self.nvectors}')
-        a = self.bits[position, :, vector // 8]
-        m = self.mask[vector % 8]
-        if self.vdim == 1:
-            self._set_value_vd1(a, m, v)
-        elif self.vdim == 2:
-            self._set_value_vd2(a, m, v)
-        elif self.vdim == 3:
-            self._set_value_vd3(a, m, v)
-    
-    def set_values(self, vector, v, mapping=None, inversions=None):
-        if vector >= self.nvectors:
-            raise IndexError(f'vector out of range: {vector} >= {self.nvectors}')
-        if not mapping:
-            mapping = [y for y in range(len(v))]
-        if inversions is None:
-            inversions = [False] * len(v)
-        for i, c in enumerate(v):
-            if inversions[i]:
-                if c == '1':
-                    c = '0'
-                elif c == '0':
-                    c = '1'
-                elif c == 'H':
-                    c = 'L'
-                elif c == 'L':
-                    c = 'H'
-                elif c == 'R':
-                    c = 'F'
-                elif c == 'F':
-                    c = 'R'
-            self.set_value(vector, mapping[i], c)
-    
-    def set_values_for_position(self, position, values):
-        for i, v in enumerate(values):
-            self.set_value(i, position, v)
-            
-    def __setitem__(self, vector, value):
-        for i, c in enumerate(value):
-            self.set_value(vector, i, c)
-
-    def __getitem__(self, vector):
-        if isinstance(vector, slice):
-            first = self.get_values_for_position(0)[vector]
-            ret = PackedVectors(len(first), self.width, self.vdim)
-            ret.set_values_for_position(0, first)
-            for pos in range(1, self.width):
-                ret.set_values_for_position(pos, self.get_values_for_position(pos)[vector])
-            return ret
-        return ''.join(self.get_value(vector, pos) for pos in range(len(self.bits)))
-
-    @staticmethod
-    def _set_value_vd1(a, m, v):
-        if v in [True, 1, '1', 'H', 'h']:
-            a[0] |= m[1]
-        else:
-            a[0] &= m[0]
-    
-    @staticmethod
-    def _set_value_vd2(a, m, v):
-        if v in [True, 1, '1', 'H', 'h']:
-            a[0] |= m[1]
-            a[1] |= m[1]
-        elif v in [False, 0, '0', 'L', 'l']:
-            a[0] &= m[0]
-            a[1] |= m[1]
-        elif v in ['X', 'x']:
-            a[0] |= m[1]
-            a[1] &= m[0]
-        else:
-            a[0] &= m[0]
-            a[1] &= m[0]
-
-    #   i fb act
-    # a 0 1 2
-    # - 0 0 0  None, '-'
-    # 0 0 1 0  False, 0, '0', 'l', 'L'
-    # 1 1 0 0  True, 1, '1', 'h', 'H'
-    # X 1 1 0  'x', 'X'
-    # / 0 0 1  '/', 'r', 'R'
-    # ^ 0 1 1  '^', 'p', 'P'
-    # v 1 0 1  'v', 'n', 'N'
-    # \ 1 1 1  '\', 'f', 'F'
-    @staticmethod
-    def _set_value_vd3(a, m, v):
-        if v in [False, 0, '0', 'L', 'l']:
-            a[0] &= m[0]
-            a[1] |= m[1]
-            a[2] &= m[0]
-        elif v in [True, 1, '1', 'H', 'h']:
-            a[0] |= m[1]
-            a[1] &= m[0]
-            a[2] &= m[0]
-        elif v in ['X', 'x']:
-            a[0] |= m[1]
-            a[1] |= m[1]
-            a[2] &= m[0]
-        elif v in ['/', 'r', 'R']:
-            a[0] &= m[0]
-            a[1] &= m[0]
-            a[2] |= m[1]
-        elif v in ['^', 'p', 'P']:
-            a[0] &= m[0]
-            a[1] |= m[1]
-            a[2] |= m[1]
-        elif v in ['v', 'n', 'N']:
-            a[0] |= m[1]
-            a[1] &= m[0]
-            a[2] |= m[1]
-        elif v in ['\\', 'f', 'F']:
-            a[0] |= m[1]
-            a[1] |= m[1]
-            a[2] |= m[1]
-        else:
-            a[0] &= m[0]
-            a[1] &= m[0]
-            a[2] &= m[0]
-                                    
-    def __repr__(self):
-        return f'<PackedVectors nvectors={self.nvectors}, width={self.width}, vdim={self.vdim}>'
-
-    def __str__(self):
-        lst = []
-        for p in range(self.nvectors):
-            lst.append(''.join(self.get_value(p, w) for w in range(len(self.bits))))
-        if len(lst) == 0: return ''
-        if len(lst[0]) > 64:
-            lst = [s[:32] + '...' + s[-32:] for s in lst]
-        if len(lst) <= 16:
-            return '\n'.join(lst)
-        else:
-            return '\n'.join(lst[:8]) + '\n...\n' + '\n'.join(lst[-8:])
-            
-    def diff(self, other, out=None):
-        if out is None:
-            out = np.zeros((self.width, self.bits.shape[-1]), dtype='uint8')
-        out[...] = (self.value_bits ^ other.value_bits) & self.care_bits & other.care_bits
-        return out

src/kyupy/saed.py

@@ -1,289 +0,0 @@
-from kyupy.circuit import Node, Line
-
-
-def pin_index(cell_type, pin):
-    if cell_type.startswith('HADD') and pin == 'B0': return 1
-    if cell_type.startswith('HADD') and pin == 'SO': return 1
-    if cell_type.startswith('MUX21') and pin == 'S': return 2
-    if cell_type.startswith('SDFF') and pin == 'QN': return 1
-    if cell_type.startswith('DFF') and pin == 'QN': return 1
-    if cell_type.startswith('DFF') and pin == 'CLK': return 1
-    if cell_type.startswith('DFF') and pin == 'RSTB': return 2
-    if cell_type.startswith('DFF') and pin == 'SETB': return 3
-    if pin in ['A2', 'IN2', 'SE', 'B', 'CO']: return 1
-    if pin in ['A3', 'IN3', 'SI', 'CI']: return 2
-    if pin == 'A4' or pin == 'IN4' or pin == 'CLK': return 3  # CLK for scan cells SDFF
-    if pin == 'A5' or pin == 'IN5' or pin == 'RSTB': return 4
-    if pin == 'A6' or pin == 'IN6' or pin == 'SETB': return 5
-    return 0
-
-
-def pin_is_output(kind, pin):
-    if 'MUX' in kind and pin == 'S':
-        return False
-    return pin in ['Q', 'QN', 'Z', 'ZN', 'Y', 'CO', 'S', 'SO', 'C1']
-
-
-def add_and_connect(circuit, name, kind, in1=None, in2=None, out=None):
-    n = Node(circuit, name, kind)
-    if in1 is not None:
-        n.ins[0] = in1
-        in1.reader = n
-        in1.reader_pin = 0
-    if in2 is not None:
-        n.ins[1] = in2
-        in2.reader = n
-        in2.reader_pin = 1
-    if out is not None:
-        n.outs[0] = out
-        out.driver = n
-        out.driver_pin = 0
-    return n
-
-
-def split_complex_gates(circuit):
-    node_list = circuit.nodes
-    for n in node_list:
-        name = n.name
-        ins = n.ins
-        outs = n.outs
-        if n.kind.startswith('AO21X'):
-            n.remove()
-            n_and = add_and_connect(circuit, name+'~and', 'AND2', ins[0], ins[1], None)
-            n_or = add_and_connect(circuit, name+'~or', 'OR2', None, ins[2], outs[0])
-            Line(circuit, n_and, n_or)
-        elif n.kind.startswith('AOI21X'):
-            n.remove()
-            n_and = add_and_connect(circuit, name+'~and', 'AND2', ins[0], ins[1], None)
-            n_nor = add_and_connect(circuit, name+'~nor', 'NOR2', None, ins[2], outs[0])
-            Line(circuit, n_and, n_nor)
-        elif n.kind.startswith('OA21X'):
-            n.remove()
-            n_or = add_and_connect(circuit, name+'~or', 'OR2', ins[0], ins[1], None)
-            n_and = add_and_connect(circuit, name+'~and', 'AND2', None, ins[2], outs[0])
-            Line(circuit, n_or, n_and)
-        elif n.kind.startswith('OAI21X'):
-            n.remove()
-            n_or = add_and_connect(circuit, name+'~or', 'OR2', ins[0], ins[1], None)
-            n_nand = add_and_connect(circuit, name+'~nand', 'NAND2', None, ins[2], outs[0])
-            Line(circuit, n_or, n_nand)
-        elif n.kind.startswith('OA22X'):
-            n.remove()
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n_or1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n_and = add_and_connect(circuit, name+'~and', 'AND2', None, None, outs[0])
-            Line(circuit, n_or0, n_and)
-            Line(circuit, n_or1, n_and)
-        elif n.kind.startswith('OAI22X'):
-            n.remove()
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n_or1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n_nand = add_and_connect(circuit, name+'~nand', 'NAND2', None, None, outs[0])
-            Line(circuit, n_or0, n_nand)
-            Line(circuit, n_or1, n_nand)
-        elif n.kind.startswith('AO22X'):
-            n.remove()
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n_or = add_and_connect(circuit, name+'~or', 'OR2', None, None, outs[0])
-            Line(circuit, n_and0, n_or)
-            Line(circuit, n_and1, n_or)
-        elif n.kind.startswith('AOI22X'):
-            n.remove()
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n_nor = add_and_connect(circuit, name+'~nor', 'NOR2', None, None, outs[0])
-            Line(circuit, n_and0, n_nor)
-            Line(circuit, n_and1, n_nor)
-        elif n.kind.startswith('AO221X'):
-            n.remove()
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', None, None, None)
-            n_or1 = add_and_connect(circuit, name+'~or1', 'OR2', None, ins[4], outs[0])
-            Line(circuit, n_and0, n_or0)
-            Line(circuit, n_and1, n_or0)
-            Line(circuit, n_or0, n_or1)     
-        elif n.kind.startswith('AOI221X'):
-            n.remove()
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n_or = add_and_connect(circuit, name+'~or', 'OR2', None, None, None)
-            n_nor = add_and_connect(circuit, name+'~nor', 'NOR2', None, ins[4], outs[0])
-            Line(circuit, n_and0, n_or)
-            Line(circuit, n_and1, n_or)
-            Line(circuit, n_or, n_nor)     
-        elif n.kind.startswith('OA221X'):
-            n.remove()
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n_or1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', None, None, None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', None, ins[4], outs[0])
-            Line(circuit, n_or0, n_and0)
-            Line(circuit, n_or1, n_and0)
-            Line(circuit, n_and0, n_and1)    
-        elif n.kind.startswith('OAI221X'):
-            n.remove()
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n_or1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', None, None, None)
-            n_nand1 = add_and_connect(circuit, name+'~nand1', 'NAND2', None, ins[4], outs[0])
-            Line(circuit, n_or0, n_and0)
-            Line(circuit, n_or1, n_and0)
-            Line(circuit, n_and0, n_nand1)
-        elif n.kind.startswith('AO222X'):
-            n.remove()
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n_and2 = add_and_connect(circuit, name+'~and2', 'AND2', ins[4], ins[5], None)
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', None, None, None)
-            n_or1 = add_and_connect(circuit, name+'~or1', 'OR2', None, None, outs[0])
-            Line(circuit, n_and0, n_or0)
-            Line(circuit, n_and1, n_or0)
-            Line(circuit, n_and2, n_or1)
-            Line(circuit, n_or0, n_or1)
-        elif n.kind.startswith('AOI222X'):
-            n.remove()
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n_and2 = add_and_connect(circuit, name+'~and2', 'AND2', ins[4], ins[5], None)
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', None, None, None)
-            n_nor1 = add_and_connect(circuit, name+'~nor1', 'NOR2', None, None, outs[0])
-            Line(circuit, n_and0, n_or0)
-            Line(circuit, n_and1, n_or0)
-            Line(circuit, n_and2, n_nor1)
-            Line(circuit, n_or0, n_nor1)
-        elif n.kind.startswith('OA222X'):
-            n.remove()
-            n_or0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n_or1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n_or2 = add_and_connect(circuit, name+'~or2', 'OR2', ins[4], ins[5], None)
-            n_and0 = add_and_connect(circuit, name+'~and0', 'AND2', None, None, None)
-            n_and1 = add_and_connect(circuit, name+'~and1', 'AND2', None, None, outs[0])
-            Line(circuit, n_or0, n_and0)
-            Line(circuit, n_or1, n_and0)
-            Line(circuit, n_or2, n_and1)
-            Line(circuit, n_and0, n_and1)
-        elif n.kind.startswith('OAI222X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n2 = add_and_connect(circuit, name+'~or2', 'OR2', ins[4], ins[5], None)
-            n3 = add_and_connect(circuit, name+'~and0', 'AND2', None, None, None)
-            n4 = add_and_connect(circuit, name+'~nand1', 'NAND2', None, None, outs[0])
-            Line(circuit, n0, n3)
-            Line(circuit, n1, n3)
-            Line(circuit, n2, n4)
-            Line(circuit, n3, n4)
-        elif n.kind.startswith('AND3X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~and1', 'AND2', None, ins[2], outs[0])
-            Line(circuit, n0, n1)
-        elif n.kind.startswith('OR3X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~or1', 'OR2', None, ins[2], outs[0])
-            Line(circuit, n0, n1)
-        elif n.kind.startswith('XOR3X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~xor0', 'XOR2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~xor1', 'XOR2', None, ins[2], outs[0])
-            Line(circuit, n0, n1)
-        elif n.kind.startswith('NAND3X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~and', 'AND2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~nand', 'NAND2', None, ins[2], outs[0])
-            Line(circuit, n0, n1)
-        elif n.kind.startswith('NOR3X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~or', 'OR2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~nor', 'NOR2', None, ins[2], outs[0])
-            Line(circuit, n0, n1)
-        elif n.kind.startswith('XNOR3X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~xor', 'XOR2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~xnor', 'XNOR2', None, ins[2], outs[0])
-            Line(circuit, n0, n1)
-        elif n.kind.startswith('AND4X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n2 = add_and_connect(circuit, name+'~and2', 'AND2', None, None, outs[0])
-            Line(circuit, n0, n2)
-            Line(circuit, n1, n2)
-        elif n.kind.startswith('OR4X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n2 = add_and_connect(circuit, name+'~or2', 'OR2', None, None, outs[0])
-            Line(circuit, n0, n2)
-            Line(circuit, n1, n2)
-        elif n.kind.startswith('NAND4X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~and0', 'AND2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~and1', 'AND2', ins[2], ins[3], None)
-            n2 = add_and_connect(circuit, name+'~nand2', 'NAND2', None, None, outs[0])
-            Line(circuit, n0, n2)
-            Line(circuit, n1, n2)
-        elif n.kind.startswith('NOR4X'):
-            n.remove()
-            n0 = add_and_connect(circuit, name+'~or0', 'OR2', ins[0], ins[1], None)
-            n1 = add_and_connect(circuit, name+'~or1', 'OR2', ins[2], ins[3], None)
-            n2 = add_and_connect(circuit, name+'~nor2', 'NOR2', None, None, outs[0])
-            Line(circuit, n0, n2)
-            Line(circuit, n1, n2)
-        elif n.kind.startswith('FADDX'):
-            n.remove()
-            # forks for fan-outs
-            f_a = add_and_connect(circuit, name + '~fork0', '__fork__', ins[0])
-            f_b = add_and_connect(circuit, name + '~fork1', '__fork__', ins[1])
-            f_ci = add_and_connect(circuit, name + '~fork2', '__fork__', ins[2])
-            f_ab = Node(circuit, name + '~fork3')
-            # sum-block
-            n_xor0 = Node(circuit, name + '~xor0', 'XOR2')
-            Line(circuit, f_a, n_xor0)
-            Line(circuit, f_b, n_xor0)
-            Line(circuit, n_xor0, f_ab)
-            if len(outs) > 0 and outs[0] is not None:
-                n_xor1 = add_and_connect(circuit, name + '~xor1', 'XOR2', None, None, outs[0])
-                Line(circuit, f_ab, n_xor1)
-                Line(circuit, f_ci, n_xor1)
-            # carry-block
-            if len(outs) > 1 and outs[1] is not None:
-                n_and0 = Node(circuit, name + '~and0', 'AND2')
-                Line(circuit, f_ab, n_and0)
-                Line(circuit, f_ci, n_and0)
-                n_and1 = Node(circuit, name + '~and1', 'AND2')
-                Line(circuit, f_a, n_and1)
-                Line(circuit, f_b, n_and1)
-                n_or = add_and_connect(circuit, name + '~or0', 'OR2', None, None, outs[1])
-                Line(circuit, n_and0, n_or)
-                Line(circuit, n_and1, n_or)
-        elif n.kind.startswith('HADDX'):
-            n.remove()
-            # forks for fan-outs
-            f_a = add_and_connect(circuit, name + '~fork0', '__fork__', ins[0])
-            f_b = add_and_connect(circuit, name + '~fork1', '__fork__', ins[1])
-            n_xor0 = add_and_connect(circuit, name + '~xor0', 'XOR2', None, None, outs[1])
-            Line(circuit, f_a, n_xor0)
-            Line(circuit, f_b, n_xor0)
-            n_and0 = add_and_connect(circuit, name + '~and0', 'AND2', None, None, outs[0])
-            Line(circuit, f_a, n_and0)
-            Line(circuit, f_b, n_and0)
-        elif n.kind.startswith('MUX21X'):
-            n.remove()
-            f_s = add_and_connect(circuit, name + '~fork0', '__fork__', ins[2])
-            n_not = Node(circuit, name + '~not', 'INV')
-            Line(circuit, f_s, n_not)
-            n_and0 = add_and_connect(circuit, name + '~and0', 'AND2', ins[0])
-            n_and1 = add_and_connect(circuit, name + '~and1', 'AND2', ins[1])
-            n_or0 = add_and_connect(circuit, name + '~or0', 'OR2', None, None, outs[0])
-            Line(circuit, n_not, n_and0)
-            Line(circuit, f_s, n_and1)
-            Line(circuit, n_and0, n_or0)
-            Line(circuit, n_and1, n_or0)
-        elif n.kind.startswith('DFFSSR'):
-            n.kind = 'DFFX1'
-            n_and0 = add_and_connect(circuit, name + '~and0', 'AND2', ins[0], ins[2], None)
-            Line(circuit, n_and0, (n, 0))

src/kyupy/sdf.py

@@ -1,145 +1,153 @@
+"""A simple and incomplete parser for the Standard Delay Format (SDF).
+
+This parser extracts pin-to-pin delay and interconnect delay information from SDF files.
+Sophisticated timing specifications (timing checks, conditional delays, etc.) are ignored.
+
+The functions :py:func:`parse` and :py:func:`load` return an intermediate representation (:class:`DelayFile` object).
+Call :py:func:`DelayFile.iopaths` and :py:func:`DelayFile.interconnects` to generate delay information for a given circuit.
+"""
+
+from collections import namedtuple
+import re
+
 import numpy as np
 from lark import Lark, Transformer
-from collections import namedtuple
-from . import log
-import gzip
+
+from . import log, readtext
+from .circuit import Circuit
+from .techlib import TechLib
+
 
 Interconnect = namedtuple('Interconnect', ['orig', 'dest', 'r', 'f'])
 IOPath = namedtuple('IOPath', ['ipin', 'opin', 'r', 'f'])
 
 
 class DelayFile:
+    """An intermediate representation of an SDF file.
+    """
     def __init__(self, name, cells):
         self.name = name
-        if None in cells:
-            self.interconnects = cells[None]
-        else:
-            self.interconnects = None
+        self._interconnects = cells.get(None, None)
         self.cells = dict((n, l) for n, l in cells.items() if n)
 
     def __repr__(self):
         return '\n'.join(f'{n}: {l}' for n, l in self.cells.items()) + '\n' + \
-               '\n'.join(str(i) for i in self.interconnects)
+               '\n'.join(str(i) for i in self._interconnects)
 
-    def annotation(self, circuit, pin_index_f, dataset=1, interconnect=True, ffdelays=True):
+    def iopaths(self, circuit:Circuit, tlib:TechLib):
+        """Constructs an ndarray containing all IOPATH delays.
+
+        All IOPATH delays for a node ``n`` are annotated to the line connected to the input pin specified in the IOPATH.
+
+        Limited support of SDF spec:
+
+        * Only ABSOLUTE delay values are supported.
+        * Only two delvals per delval_list is supported. First is rising/posedge, second is falling/negedge
+          transition at the output of the IOPATH (SDF spec, pp. 3-17).
+        * PATHPULSE declarations are ignored.
+
+        The axes convention of KyuPy's delay data arrays is as follows:
+
+        * Axis 0: dataset (usually 3 datasets per SDF-file)
+        * Axis 1: line index (e.g. ``n.ins[0]``, ``n.ins[1]``)
+        * Axis 2: polarity of the transition at the IOPATH-input (e.g. at ``n.ins[0]`` or ``n.ins[1]``), 0='rising/posedge', 1='falling/negedge'
+        * Axis 3: polarity of the transition at the IOPATH-output (at ``n.outs[0]``), 0='rising/posedge', 1='falling/negedge'
         """
-        Constructs an 3-dimensional array with timing data for each line in `circuit`.
-        Dimension 1 of the returned array is the line index.
-        Dimension 2 is the type of timing data: 0:`delay`, 1:`pulse rejection limit`.
-        Dimension 3 is the polarity at the output of the reading node: 0:`rising`, 1:`falling`.
 
-        The polarity for pulse rejection is determined by the latter transition of the pulse.
-        E.g., timing[42,1,0] is the rejection limit of a negative pulse at the output of the reader of line 42.
+        def find_cell(name:str):
+            if name not in circuit.cells: name = name.replace('\\', '')
+            if name not in circuit.cells: name = name.replace('[', '_').replace(']', '_')
+            return circuit.cells.get(name, None)
 
-        An IOPATH delay for a node is annotated to the line connected to the input pin specified in the IOPATH.
+        delays = np.zeros((len(circuit.lines), 2, 2, 3), dtype=np.float32)  # dataset last during construction.
 
-        Currently, only ABSOLUTE IOPATH and INTERCONNECT delays are supported.
-        Pulse rejection limits are derived from absolute delays, explicit declarations (PATHPULSE etc.) are ignored.
-
-
-        :param ffdelays:
-        :param interconnect:
-        :param pin_index_f:
-        :param circuit:
-        :type dataset: int or tuple
-        """
-        def select_del(_delvals, idx):
-            if type(dataset) is tuple:
-                s = 0
-                for d in dataset:
-                    s += _delvals[idx][d]
-                return s / len(dataset)
-            else:
-                return _delvals[idx][dataset]
-        
-        def find_cell(name):
-            if name not in circuit.cells:
+        with log.limit(50):
+            for name, iopaths in self.cells.items():
                 name = name.replace('\\', '')
-            if name not in circuit.cells:
-                name = name.replace('[', '_').replace(']', '_')
-            if name not in circuit.cells:
-                return None
-            return circuit.cells[name]
-        
-        timing = np.zeros((len(circuit.lines), 2, 2))
-        for cn, iopaths in self.cells.items():
-            for ipn, opn, *delvals in iopaths:
-                delvals = [d if len(d) > 0 else [0, 0, 0] for d in delvals]
-                if max(max(delvals)) == 0:
-                    continue
-                cell = find_cell(cn)
-                if cell is None:
-                    log.warn(f'Cell from SDF not found in circuit: {cn}')
-                    continue
-                ipin = pin_index_f(cell.kind, ipn)
-                opin = pin_index_f(cell.kind, opn)
-                kind = cell.kind.lower()
-
-                ipn2 = ipn.replace('(posedge A1)', 'A1').replace('(negedge A1)', 'A1')\
-                    .replace('(posedge A2)', 'A2').replace('(negedge A2)', 'A2')
-                
-                def add_delays(_line):
-                    if _line is not None:
-                        timing[_line.index, :, 0] += select_del(delvals, 0)
-                        timing[_line.index, :, 1] += select_del(delvals, 1)
-
-                take_avg = False
-                if kind.startswith('sdff'):
-                    if not ipn.startswith('(posedge CLK'):
-                        continue
-                    if ffdelays and (len(cell.outs) > opin):
-                        add_delays(cell.outs[opin])
+                if cell := circuit.cells.get(name, None):
+                    for i_pin_spec, o_pin_spec, *dels in iopaths:
+                        if i_pin_spec.startswith('(posedge '): i_pol_idxs = [0]
+                        elif i_pin_spec.startswith('(negedge '): i_pol_idxs = [1]
+                        else: i_pol_idxs = [0, 1]
+                        i_pin_spec = re.sub(r'\((neg|pos)edge ([^)]+)\)', r'\2', i_pin_spec)
+                        if line := cell.ins[tlib.pin_index(cell.kind, i_pin_spec)]:
+                            delays[line, i_pol_idxs] = [d if len(d) > 0 else [0, 0, 0] for d in dels]
+                        else:
+                            log.warn(f'No line to annotate in circuit: {i_pin_spec} for {cell}')
                 else:
-                    if kind.startswith(('xor', 'xnor')):
-                        ipin = pin_index_f(cell.kind, ipn2)
-                        # print(ipn, ipin, times[cell.i_lines[ipin].index, 0, 0])
-                        take_avg = timing[cell.ins[ipin].index].sum() > 0
-                    add_delays(cell.ins[ipin])
-                    if take_avg:
-                        timing[cell.ins[ipin].index] /= 2
-        
-        if not interconnect or self.interconnects is None:
-            return timing
-        
-        for n1, n2, *delvals in self.interconnects:
+                    log.warn(f'Name from SDF not found in circuit: {name}')
+
+        return np.moveaxis(delays, -1, 0)
+
+    def interconnects(self, circuit:Circuit, tlib:TechLib):
+        """Constructs an ndarray containing all INTERCONNECT delays.
+
+        To properly annotate interconnect delays, the circuit model has to include a '__fork__' node on
+        every signal and every fanout-branch. The Verilog parser aids in this by setting the parameter
+        `branchforks=True` in :py:func:`~kyupy.verilog.parse` or :py:func:`~kyupy.verilog.load`.
+
+        Limited support of SDF spec:
+
+        * Only ABSOLUTE delay values are supported.
+        * Only two delvals per delval_list is supported. First is rising/posedge, second is falling/negedge
+          transition.
+        * PATHPULSE declarations are ignored.
+
+        The axes convention of KyuPy's delay data arrays is as follows:
+
+        * Axis 0: dataset (usually 3 datasets per SDF-file)
+        * Axis 1: line index. Usually input line of a __fork__.
+        * Axis 2: (axis of size 2 for compatability to IOPATH results. Values are broadcast along this axis.)
+        * Axis 3: polarity of the transition, 0='rising/posedge', 1='falling/negedge'
+        """
+
+        delays = np.zeros((len(circuit.lines), 2, 2, 3), dtype=np.float32)  # dataset last during construction.
+
+        def split_node_port(n_str):
+            if (split := n_str.rfind('.')) > 0 or (split := n_str.rfind('/')) > 0:
+                return n_str[:split], n_str[split+1:]
+            return (n_str, None)
+
+        nonfork_annotations = 0
+        for n1, n2, *delvals in self._interconnects:
             delvals = [d if len(d) > 0 else [0, 0, 0] for d in delvals]
-            if max(max(delvals)) == 0:
+            if max(max(delvals)) == 0: continue
+            cn1, pn1 = split_node_port(n1)
+            cn2, pn2 = split_node_port(n2)
+            cn1 = cn1.replace('\\','')
+            cn2 = cn2.replace('\\','')
+            c1, c2 = circuit.cells[cn1], circuit.cells[cn2]
+            p1 = tlib.pin_index(c1.kind, pn1) if pn1 is not None else 0
+            p2 = tlib.pin_index(c2.kind, pn2) if pn2 is not None else 0
+            if len(c1.outs) <= p1 or c1.outs[p1] is None:
+                log.warn(f'No line to annotate pin {pn1} of {c1}')
                 continue
-            if '/' in n1:
-                i = n1.rfind('/')
-                cn1 = n1[0:i]
-                pn1 = n1[i+1:]
-            else:
-                cn1, pn1 = (n1, 'Z')
-            if '/' in n2:
-                i = n2.rfind('/')
-                cn2 = n2[0:i]
-                pn2 = n2[i+1:]
-            else:
-                cn2, pn2 = (n2, 'IN')
-            c1 = find_cell(cn1)
-            if c1 is None:
-                log.warn(f'Cell from SDF not found in circuit: {cn1}')
+            if len(c2.ins) <= p2 or c2.ins[p2] is None:
+                log.warn(f'No line to annotate pin {pn2} of {c2}')
                 continue
-            c2 = find_cell(cn2)
-            if c2 is None:
-                log.warn(f'Cell from SDF not found in circuit: {cn2}')
-                continue
-            p1, p2 = pin_index_f(c1.kind, pn1), pin_index_f(c2.kind, pn2)
-            line = None
-            f1, f2 = c1.outs[p1].reader, c2.ins[p2].driver
-            if f1 != f2:  # possible branchfork
-                assert len(f2.ins) == 1
-                line = f2.ins[0]
-                assert f1.outs[f2.ins[0].driver_pin] == line
-            elif len(f2.outs) == 1:  # no fanout?
-                line = f2.ins[0]
-            if line is not None:
-                timing[line.index, :, 0] += select_del(delvals, 0)
-                timing[line.index, :, 1] += select_del(delvals, 1)
+            f1, f2 = c1.outs[p1].reader, c2.ins[p2].driver  # find the forks between cells.
+            if f1 == c2 and f2 == c1:
+                nonfork_annotations += 1
+                if nonfork_annotations < 10:
+                    log.warn(f'No fork between {c1.name}/{p1} and {c2.name}/{p2}, using {c2.name}/{p2}')
+                line = c2.ins[p2]
             else:
-                log.warn(f'No branchfork for annotating interconnect delay {c1.name}/{p1}->{c2.name}/{p2}')
-        return timing
+                assert f1.kind == '__fork__'
+                assert f2.kind == '__fork__'
+                if len(f2.outs) == 1:
+                    assert f1 == f2 or f1.outs[f2.ins[0].driver_pin] == f2.ins[0]
+                    line = f2.ins[0]
+                else:
+                    nonfork_annotations += 1
+                    if nonfork_annotations < 10:
+                        log.warn(f'No branchfork between {c1.name}/{p1} and {c2.name}/{p2}, using {c2.name}/{p2}')
+                    line = c2.ins[p2]
+            delays[line, :] = delvals
+
+        if nonfork_annotations > 0:
+            log.warn(f'{nonfork_annotations} interconnect annotations were moved to gate inputs due to missing forks.')
+
+        return np.moveaxis(delays, -1, 0)
 
 
 def sanitize(args):
@@ -163,6 +171,10 @@ class SdfTransformer(Transformer):
         entries = [e for a in args if hasattr(a, 'children') for e in a.children]
         return name, entries
 
+    @staticmethod
+    def cond(args):  # ignore conditions
+        return args[1]
+
     @staticmethod
     def start(args):
         name = next((a for a in args if isinstance(a, str)), None)
@@ -170,8 +182,7 @@ class SdfTransformer(Transformer):
         return DelayFile(name, cells)
 
 
-def parse(sdf):
-    grammar = r"""
+GRAMMAR = r"""
     start: "(DELAYFILE" ( "(SDFVERSION" _NOB ")"
         | "(DESIGN" "\"" NAME "\"" ")"
         | "(DATE" _NOB ")"
@@ -188,9 +199,12 @@ def parse(sdf):
         | "(INSTANCE" ID? ")"
         | "(TIMINGCHECK" _ignore* ")"
         | delay )* ")"
-    delay: "(DELAY" "(ABSOLUTE" (interconnect | iopath)* ")" ")"
+    delay: "(DELAY" "(ABSOLUTE" (interconnect | iopath | cond)* ")" ")"
     interconnect: "(INTERCONNECT" ID ID triple* ")"
     iopath: "(IOPATH" ID_OR_EDGE ID_OR_EDGE triple* ")"
+    cond: "(" "COND" cond_port_expr iopath ")"
+    ?cond_port_expr: ID | "(" cond_port_expr ")" | cond_port_expr BINARY_OP cond_port_expr
+    BINARY_OP: /&&/ | /==/
     NAME: /[^"]+/
     ID_OR_EDGE: ( /[^() ]+/ | "(" /[^)]+/ ")" )
     ID: ( /[^"() ]+/ | "\"" /[^"]+/ "\"" )
@@ -201,13 +215,16 @@ def parse(sdf):
     %ignore ( /\r?\n/ | COMMENT )+
     %ignore /[\t\f ]+/
     """
-    if '\n' not in str(sdf):  # One line?: Assuming it is a file name.
-        if str(sdf).endswith('.gz'):
-            with gzip.open(sdf, 'rt') as f:
-                text = f.read()
-        else:
-            with open(sdf, 'r') as f:
-                text = f.read()
-    else:
-        text = str(sdf)
-    return Lark(grammar, parser="lalr", transformer=SdfTransformer()).parse(text)
+
+
+def parse(text):
+    """Parses the given ``text`` and returns a :class:`DelayFile` object."""
+    return Lark(GRAMMAR, parser="lalr", transformer=SdfTransformer()).parse(text)
+
+
+def load(file) -> DelayFile:
+    """Parses the contents of ``file`` and returns a :class:`DelayFile` object.
+
+    Files with `.gz`-suffix are decompressed on-the-fly.
+    """
+    return parse(readtext(file)) # type: ignore

src/kyupy/sim.py

@@ -0,0 +1,367 @@
+
+from collections import defaultdict
+from bisect import bisect, insort_left
+
+import numpy as np
+
+from . import log
+from .circuit import Circuit
+from .techlib import KYUPY
+
+BUF1 = np.uint16(0b1010_1010_1010_1010)
+INV1 = ~BUF1
+
+__const0__ = BUF1
+__const1__ = INV1
+
+AND2 = np.uint16(0b1000_1000_1000_1000)
+AND3 = np.uint16(0b1000_0000_1000_0000)
+AND4 = np.uint16(0b1000_0000_0000_0000)
+
+NAND2, NAND3, NAND4 = ~AND2, ~AND3, ~AND4
+
+OR2 = np.uint16(0b1110_1110_1110_1110)
+OR3 = np.uint16(0b1111_1110_1111_1110)
+OR4 = np.uint16(0b1111_1111_1111_1110)
+
+NOR2, NOR3, NOR4 = ~OR2, ~OR3, ~OR4
+
+XOR2 = np.uint16(0b0110_0110_0110_0110)
+XOR3 = np.uint16(0b1001_0110_1001_0110)
+XOR4 = np.uint16(0b0110_1001_1001_0110)
+
+XNOR2, XNOR3, XNOR4 = ~XOR2, ~XOR3, ~XOR4
+
+AO21 = np.uint16(0b1111_1000_1111_1000)  # (i0 & i1) | i2
+AO22 = np.uint16(0b1111_1000_1000_1000)  # (i0 & i1) | (i2 & i3)
+OA21 = np.uint16(0b1110_0000_1110_0000)  # (i0 | i1) & i2
+OA22 = np.uint16(0b1110_1110_1110_0000)  # (i0 | i1) & (i2 | i3)
+
+AOI21, AOI22, OAI21, OAI22 = ~AO21, ~AO22, ~OA21, ~OA22
+
+AO211 = np.uint16(0b1111_1111_1111_1000)  # (i0 & i1) | i2 | i3
+OA211 = np.uint16(0b1110_0000_0000_0000)  # (i0 | i1) & i2 & i3
+
+AOI211, OAI211 = ~AO211, ~OA211
+
+MUX21 = np.uint16(0b1100_1010_1100_1010)  # z = i1 if i2 else i0 (i2 is select)
+
+names = dict([(v, k) for k, v in globals().items() if isinstance(v, np.uint16) and '__' not in k])
+
+prim2name = dict([(v, k) for k, v in globals().items() if isinstance(v, np.uint16) and '__' not in k])
+name2prim = dict([(k, v) for k, v in globals().items() if isinstance(v, np.uint16)])
+
+kind_prefixes = {
+    'nand': (NAND4, NAND3, NAND2),
+    'nor': (NOR4, NOR3, NOR2),
+    'and': (AND4, AND3, AND2),
+    'or': (OR4, OR3, OR2),
+    'isolor': (OR2, OR2, OR2),
+    'xor': (XOR4, XOR3, XOR2),
+    'xnor': (XNOR4, XNOR3, XNOR2),
+
+    'not': (INV1, INV1, INV1),
+    'inv': (INV1, INV1, INV1),
+    'ibuf': (INV1, INV1, INV1),
+    '__const1__': (INV1, INV1, INV1),
+    'tieh': (INV1, INV1, INV1),
+
+    'buf': (BUF1, BUF1, BUF1),
+    'nbuf': (BUF1, BUF1, BUF1),
+    'delln': (BUF1, BUF1, BUF1),
+    '__const0__': (BUF1, BUF1, BUF1),
+    'tiel': (BUF1, BUF1, BUF1),
+
+    'ao211': (AO211, AO211, AO211),
+    'oa211': (OA211, OA211, OA211),
+    'aoi211': (AOI211, AOI211, AOI211),
+    'oai211': (OAI211, OAI211, OAI211),
+
+    'ao22': (AO22, AO22, AO22),
+    'aoi22': (AOI22, AOI22, AOI22),
+    'ao21': (AO21, AO21, AO21),
+    'aoi21': (AOI21, AOI21, AOI21),
+
+    'oa22': (OA22, OA22, OA22),
+    'oai22': (OAI22, OAI22, OAI22),
+    'oa21': (OA21, OA21, OA21),
+    'oai21': (OAI21, OAI21, OAI21),
+
+    'mux21': (MUX21, MUX21, MUX21),
+}
+
+class Heap:
+    def __init__(self):
+        self.chunks = dict()  # map start location to chunk size
+        self.released = list()  # chunks that were released
+        self.current_size = 0
+        self.max_size = 0
+
+    def alloc(self, size):
+        for idx, loc in enumerate(self.released):
+            if self.chunks[loc] == size:
+                del self.released[idx]
+                return loc
+            if self.chunks[loc] > size:  # split chunk
+                chunksize = self.chunks[loc]
+                self.chunks[loc] = size
+                self.chunks[loc + size] = chunksize - size
+                self.released[idx] = loc + size  # move released pointer: loc -> loc+size
+                return loc
+        # no previously released chunk; make new one
+        loc = self.current_size
+        self.chunks[loc] = size
+        self.current_size += size
+        self.max_size = max(self.max_size, self.current_size)
+        return loc
+
+    def free(self, loc):
+        size = self.chunks[loc]
+        if loc + size == self.current_size:  # end of managed area, remove chunk
+            del self.chunks[loc]
+            self.current_size -= size
+            # check and remove prev chunk if free
+            if len(self.released) > 0:
+                prev = self.released[-1]
+                if prev + self.chunks[prev] == self.current_size:
+                    chunksize = self.chunks[prev]
+                    del self.chunks[prev]
+                    del self.released[-1]
+                    self.current_size -= chunksize
+            return
+        released_idx = bisect(self.released, loc)
+        if released_idx < len(self.released) and loc + size == self.released[released_idx]:  # next chunk is free, merge
+            chunksize = size + self.chunks[loc + size]
+            del self.chunks[loc + size]
+            self.chunks[loc] = chunksize
+            size = self.chunks[loc]
+            self.released[released_idx] = loc
+        else:
+            insort_left(self.released, loc)  # put in a new release
+        if released_idx > 0:  # check if previous chunk is free
+            prev = self.released[released_idx - 1]
+            if prev + self.chunks[prev] == loc:  # previous chunk is adjacent to freed one, merge
+                chunksize = size + self.chunks[prev]
+                del self.chunks[loc]
+                self.chunks[prev] = chunksize
+                del self.released[released_idx]
+
+    def __repr__(self):
+        r = []
+        for loc in sorted(self.chunks.keys()):
+            size = self.chunks[loc]
+            released_idx = bisect(self.released, loc)
+            is_released = released_idx > 0 and len(self.released) > 0 and self.released[released_idx - 1] == loc
+            r.append(f'{loc:5d}: {"free" if is_released else "used"} {size}')
+        return "\n".join(r)
+
+
+class SimOps:
+    """A static scheduler that translates a Circuit into a topologically sorted list of basic logic operations (self.ops) and
+    a memory mapping (self.c_locs, self.c_caps) for use in simulators.
+
+    :param circuit: The circuit to create a schedule for. Must contain cells from KYUPY techlib only.
+    :param strip_forks: If enabled, the scheduler will not include fork nodes to safe simulation time.
+        Stripping forks will cause interconnect delay annotations of lines read by fork nodes to be ignored.
+    :param c_reuse: If enabled, memory of intermediate signal waveforms will be re-used. This greatly reduces
+        memory footprint, but intermediate signal waveforms become unaccessible after a propagation.
+    """
+    def __init__(self, circuit: Circuit, c_caps=1, c_caps_min=1, a_ctrl=None, c_reuse=False, strip_forks=False):
+        self.circuit = circuit
+        self.s_len = len(circuit.s_nodes(KYUPY))
+
+        if isinstance(c_caps, int):
+            c_caps = [c_caps] * (len(circuit.lines)+3)
+
+        if a_ctrl is None:
+            a_ctrl = np.zeros((len(circuit.lines)+3, 3), dtype=np.int32)  # add 3 for zero, tmp, tmp2
+            a_ctrl[:,0] = -1
+
+        # special locations and offsets in c_locs/c_caps
+        self.zero_idx = len(circuit.lines)
+        self.tmp_idx = self.zero_idx + 1
+        self.tmp2_idx = self.tmp_idx + 1
+        self.ppi_offset = self.tmp2_idx + 1
+        self.ppo_offset = self.ppi_offset + self.s_len
+        self.c_locs_len = self.ppo_offset + self.s_len
+
+        # ALAP-toposort the circuit into self.ops
+        levels = []
+
+        ppio2idx = dict((n, i) for i, n in enumerate(circuit.s_nodes(KYUPY)))
+        root_nodes = set([n for n in circuit.nodes if len(n.outs) == 0 or KYUPY.is_dff(n.kind)])  # start from POs (that are unused by circuit itself), PPOs, and any dangling nodes
+        readers = np.array([1 if l.reader in root_nodes else len(l.reader.outs) for l in circuit.lines], dtype=np.int32)  # for ref-counting forks
+
+        level_lines = [n.ins[0] for n in sorted(root_nodes, key=lambda n: n.index) if len(n.ins) > 0 ]
+        # FIXME: Should probably instanciate buffers for PPOs and attach DFF clocks
+
+        while len(level_lines) > 0:  # traverse the circuit level-wise back towards (P)PIs
+            level_ops = []
+            prev_level_lines = []
+
+            for l in level_lines:
+                n = l.driver
+                assert len(n.ins) <= 4, f'Too many input pins for node {n}. Map circuit to KYUPY techlib first using resolve_tlib_cells().'
+                in_idxs = [n.ins[x].index if len(n.ins) > x and n.ins[x] is not None else self.zero_idx for x in [0,1,2,3]]
+                if n in ppio2idx:
+                    in_idxs[1] = self.zero_idx
+                    if 'dff' in n.kind.lower() or in_idxs[0] == self.zero_idx:
+                        in_idxs[0] = self.ppi_offset + ppio2idx[n]
+                    else:  # this is a PO that is also used within the circuit. Treat as normal fork with ref counting. TODO: check strip_forks=True case for this.
+                        readers[n.ins[0]] -= 1
+                        if readers[n.ins[0]] == 0: prev_level_lines.append(n.ins[0])
+                    if l.driver_pin == 1 and 'dff' in n.kind.lower():  # second output of DFF is inverted
+                        level_ops.append((INV1, l.index, *in_idxs, *a_ctrl[l]))
+                    else:
+                        level_ops.append((BUF1, l.index, *in_idxs, *a_ctrl[l]))
+                elif n.kind == '__fork__':
+                    readers[n.ins[0]] -= 1
+                    if readers[n.ins[0]] == 0: prev_level_lines.append(n.ins[0])
+                    if not strip_forks: level_ops.append((BUF1, l.index, *in_idxs, *a_ctrl[l]))
+                else:
+                    prev_level_lines += n.ins
+                    sp = None
+                    kind = n.kind.lower()
+                    for prefix, prims in kind_prefixes.items():
+                        if kind.startswith(prefix):
+                            sp = prims[0]
+                            if in_idxs[3] == self.zero_idx:
+                                sp = prims[1]
+                                if in_idxs[2] == self.zero_idx:
+                                    sp = prims[2]
+                            break
+                    assert sp is not None, f'Unsupported node type {n}. Map circuit to KYUPY techlib first using resolce_tlib_cells().'
+                    level_ops.append((sp, l.index, *in_idxs, *a_ctrl[l]))
+
+            if len(level_ops) > 0: levels.append(level_ops)
+            level_lines = prev_level_lines
+
+        self.levels = [np.asarray(lv, dtype=np.int32) for lv in levels[::-1]]
+        level_sums = np.cumsum([0]+[len(lv) for lv in self.levels], dtype=np.int32)
+        self.level_starts, self.level_stops = level_sums[:-1], level_sums[1:]
+        # op format: [kind, out0, in0, in1, in2, in3, wsa_acc_pos, wsa_rise, wsa_fall]
+        self.ops = np.vstack(self.levels)
+
+        # create a map from fanout lines to stem lines for fork stripping
+        stems = np.full(self.c_locs_len, -1, dtype=np.int32)  # default to -1: 'no fanout line'
+        if strip_forks:
+            for f in circuit.forks.values():
+                prev_line = f.ins[0]
+                while prev_line.driver.kind == '__fork__':
+                    prev_line = prev_line.driver.ins[0]
+                for ol in f.outs:
+                    if ol is not None:
+                        stems[ol] = prev_line.index
+
+        ref_count = np.zeros(self.c_locs_len, dtype=np.int32)
+
+        for op in self.ops:
+            for x in [2, 3, 4, 5]:
+                ref_count[stems[op[x]] if stems[op[x]] >= 0 else op[x]] += 1
+
+        # combinational signal allocation table. maps line and interface indices to self.c memory locations
+        self.c_locs = np.full((self.c_locs_len,), -1, dtype=np.int32)
+        self.c_caps = np.zeros((self.c_locs_len,), dtype=np.int32)
+
+        h = Heap()
+
+        # allocate and keep memory for special fields
+        self.c_locs[self.zero_idx], self.c_caps[self.zero_idx] = h.alloc(c_caps_min), c_caps_min
+        self.c_locs[self.tmp_idx], self.c_caps[self.tmp_idx] = h.alloc(c_caps_min), c_caps_min
+        self.c_locs[self.tmp2_idx], self.c_caps[self.tmp2_idx] = h.alloc(c_caps_min), c_caps_min
+        ref_count[self.zero_idx] += 1
+        ref_count[self.tmp_idx] += 1
+        ref_count[self.tmp2_idx] += 1
+
+        # allocate and keep memory for PI/PPI, keep memory for PO/PPO (allocated later)
+        for i, n in enumerate(circuit.s_nodes(KYUPY)):
+            if 'dff' in n.kind.lower() or len(n.ins) == 0:  # PPI or PI
+                self.c_locs[self.ppi_offset + i], self.c_caps[self.ppi_offset + i] = h.alloc(c_caps_min), c_caps_min
+                ref_count[self.ppi_offset + i] += 1
+            if len(n.ins) > 0:  # PO
+                i0_idx = stems[n.ins[0]] if stems[n.ins[0]] >= 0 else n.ins[0]
+                ref_count[i0_idx] += 1
+
+        # allocate memory for the rest of the circuit
+        for ops in self.levels:
+            free_set = set()
+            for op in ops:
+                # if we fork-strip, always take the stems
+                i0_idx = stems[op[2]] if stems[op[2]] >= 0 else op[2]
+                i1_idx = stems[op[3]] if stems[op[3]] >= 0 else op[3]
+                i2_idx = stems[op[4]] if stems[op[4]] >= 0 else op[4]
+                i3_idx = stems[op[5]] if stems[op[5]] >= 0 else op[5]
+                ref_count[i0_idx] -= 1
+                ref_count[i1_idx] -= 1
+                ref_count[i2_idx] -= 1
+                ref_count[i3_idx] -= 1
+                if ref_count[i0_idx] <= 0: free_set.add(self.c_locs[i0_idx])
+                if ref_count[i1_idx] <= 0: free_set.add(self.c_locs[i1_idx])
+                if ref_count[i2_idx] <= 0: free_set.add(self.c_locs[i2_idx])
+                if ref_count[i3_idx] <= 0: free_set.add(self.c_locs[i3_idx])
+                o_idx = op[1]
+                cap = max(c_caps_min, c_caps[o_idx])
+                self.c_locs[o_idx], self.c_caps[o_idx] = h.alloc(cap), cap
+            if c_reuse:
+                for loc in free_set:
+                    if loc >= 0:  # DFF clocks are not allocated. Ignore for now.
+                        h.free(loc)
+
+        # copy memory location and capacity from stems to fanout lines
+        for lidx, stem in enumerate(stems):
+            if stem >= 0:  # if at a fanout line
+                self.c_locs[lidx], self.c_caps[lidx] = self.c_locs[stem], self.c_caps[stem]
+
+        # copy memory location to PO/PPO area
+        for i, n in enumerate(circuit.s_nodes(KYUPY)):
+            if len(n.ins) > 0:
+                self.c_locs[self.ppo_offset + i], self.c_caps[self.ppo_offset + i] = self.c_locs[n.ins[0]], self.c_caps[n.ins[0]]
+
+        # sanity check
+        for oidx, io in zip(*np.nonzero(self.c_locs[self.ops[:, 1:6]] < 0)):
+            log.error(f'unallocated {["o0","i0","i1","i2","i3"][io]} in op {self.ops[oidx]}')
+        assert np.min(self.c_locs[self.ops[:, 1:6]]) >= 0, "Unallocated signal lines."
+
+        # line use information
+        self.line_use_start = np.full(self.c_locs_len, -1, dtype=np.int32)
+        self.line_use_stop = np.full(self.c_locs_len, len(self.levels), dtype=np.int32)
+        for i, lv in enumerate(self.levels):
+            for op in lv:
+                self.line_use_start[op[1]] = i
+                for x in [2, 3, 4, 5]:
+                    self.line_use_stop[op[x]] = i
+
+        self.c_len = h.max_size
+
+        d = defaultdict(int)
+        for op in self.ops[:,0]: d[names[op]] += 1
+        self.prim_counts = dict(d)
+
+        self.pi_s_locs = np.flatnonzero(self.c_locs[self.ppi_offset+np.arange(len(self.circuit.io_nodes))] >= 0)
+        """Locations in pattern array (axis 0) of all primary inputs (excluding flip-flops)."""
+        self.po_s_locs = np.flatnonzero(self.c_locs[self.ppo_offset+np.arange(len(self.circuit.io_nodes))] >= 0)
+        """Locations in pattern array (axis 0) of all primary outputs (excluding flip-flops)."""
+        self.ppio_s_locs = np.arange(len(self.circuit.io_nodes), self.s_len)
+        """Locations in pattern array (axis 0) of all flip-flops. Pseudo-primary inputs (PPI) and pseudo-primary outputs (PPO) share the same location."""
+
+        self.pippi_s_locs = np.concatenate([self.pi_s_locs, self.ppio_s_locs])
+        """Locations in pattern array (axis 0) of all primary and pseudo-primary inputs (input ports and flip-flops)."""
+        self.poppo_s_locs = np.concatenate([self.po_s_locs, self.ppio_s_locs])
+        """Locations in pattern array (axis 0) of all primary and pseudo-primary outputs (output ports and flip-flops)."""
+
+        self.pi_c_locs = self.c_locs[self.ppi_offset+self.pi_s_locs]
+        self.po_c_locs = self.c_locs[self.ppo_offset+self.po_s_locs]
+        self.ppi_c_locs = self.c_locs[self.ppi_offset+self.ppio_s_locs]
+        self.ppo_c_locs = self.c_locs[self.ppo_offset+self.ppio_s_locs]
+
+        self.pippi_c_locs = np.concatenate([self.pi_c_locs, self.ppi_c_locs])
+        self.poppo_c_locs = np.concatenate([self.po_c_locs, self.ppo_c_locs])
+
+    def cone_ops(self, origin_lines: set[int]):
+        active_lines = {l for l in origin_lines}
+        ops = []
+        for kind, out0, in0, in1, in2, in3, *params in self.ops:
+            if in0 in active_lines or in1 in active_lines or in2 in active_lines or in3 in active_lines or out0 in origin_lines:
+                ops.append((kind, out0, in0, in1, in2, in3, *params))
+                active_lines.add(out0)
+        return np.asarray(ops, dtype=np.int32)

src/kyupy/stil.py

@@ -1,8 +1,20 @@
-from lark import Lark, Transformer
-from collections import namedtuple
+"""A simple and incomplete parser for the Standard Test Interface Language (STIL).
+
+The main purpose of this parser is to load scan pattern sets from STIL files.
+It supports only a subset of STIL.
+
+The functions :py:func:`parse` and :py:func:`load` return an intermediate representation (:py:class:`StilFile` object).
+Call :py:func:`StilFile.tests()`, :py:func:`StilFile.tests_loc()`, or :py:func:`StilFile.responses()` to
+obtain the appropriate vector sets.
+"""
+
 import re
-import gzip
-from .packed_vectors import PackedVectors
+from collections import namedtuple
+
+import numpy as np
+from lark import Lark, Transformer
+
+from . import readtext, logic
 from .logic_sim import LogicSim
 
 
@@ -11,6 +23,8 @@ ScanPattern = namedtuple('ScanPattern', ['load', 'launch', 'capture', 'unload'])
 
 
 class StilFile:
+    """An intermediate representation of a STIL file.
+    """
     def __init__(self, version, signal_groups, scan_chains, calls):
         self.version = version
         self.signal_groups = signal_groups
@@ -21,46 +35,43 @@ class StilFile:
         self.patterns = []
         launch = {}
         capture = {}
-        load = {}
+        sload = {}
         for call in self.calls:
             if call.name == 'load_unload':
                 unload = {}
                 for so_port in self.so_ports:
                     if so_port in call.parameters:
-                        unload[so_port] = call.parameters[so_port].replace('\n', '')
-                if len(launch) > 0:
-                    self.patterns.append(ScanPattern(load, launch, capture, unload))
+                        unload[so_port] = call.parameters[so_port]
+                if len(capture) > 0:
+                    self.patterns.append(ScanPattern(sload, launch, capture, unload))
                     capture = {}
                     launch = {}
-                load = {}
+                sload = {}
                 for si_port in self.si_ports:
                     if si_port in call.parameters:
-                        load[si_port] = call.parameters[si_port].replace('\n', '')
-            if call.name.endswith('_launch') or call.name.endswith('_capture'):
-                if len(launch) == 0:
-                    launch = dict((k, v.replace('\n', '')) for k, v in call.parameters.items())
-                else:
-                    capture = dict((k, v.replace('\n', '')) for k, v in call.parameters.items())
-    
+                        sload[si_port] = call.parameters[si_port]
+            if call.name.endswith('_launch'): launch = call.parameters
+            if call.name.endswith('_capture'): capture = call.parameters
+
     def _maps(self, c):
-        interface = list(c.interface) + [n for n in c.nodes if 'DFF' in n.kind]
-        intf_pos = dict([(n.name, i) for i, n in enumerate(interface)])
+        interface = list(c.io_nodes) + [n for n in c.nodes if 'DFF' in n.kind]
+        intf_pos = dict((n.name, i) for i, n in enumerate(interface))
         pi_map = [intf_pos[n] for n in self.signal_groups['_pi']]
         po_map = [intf_pos[n] for n in self.signal_groups['_po']]
         scan_maps = {}
         scan_inversions = {}
-        for chain_name, chain in self.scan_chains.items():
+        for chain in self.scan_chains.values():
             scan_map = []
             scan_in_inversion = []
             scan_out_inversion = []
             inversion = False
             for n in chain[1:-1]:
-                if n == '!': 
+                if n == '!':
                     inversion = not inversion
                 else:
                     scan_in_inversion.append(inversion)
             scan_in_inversion = list(reversed(scan_in_inversion))
-            inversion = False             
+            inversion = False
             for n in reversed(chain[1:-1]):
                 if n == '!':
                     inversion = not inversion
@@ -69,74 +80,123 @@ class StilFile:
                     scan_out_inversion.append(inversion)
             scan_maps[chain[0]] = scan_map
             scan_maps[chain[-1]] = scan_map
-            scan_inversions[chain[0]] = scan_in_inversion
-            scan_inversions[chain[-1]] = scan_out_inversion
+            scan_inversions[chain[0]] = logic.mvarray(scan_in_inversion)[0]
+            scan_inversions[chain[-1]] = logic.mvarray(scan_out_inversion)[0]
         return interface, pi_map, po_map, scan_maps, scan_inversions
-        
-    def tests(self, c):
-        interface, pi_map, po_map, scan_maps, scan_inversions = self._maps(c)
-        tests = PackedVectors(len(self.patterns), len(interface), 2)
+
+    def tests(self, circuit):
+        """Assembles and returns a scan test pattern set for given circuit.
+
+        This function assumes a static (stuck-at fault) test.
+
+        :param circuit: The circuit to assemble the patterns for. The patterns will follow the
+            :py:attr:`~kyupy.circuit.Circuit.s_nodes` ordering of the this circuit.
+        :return: A 4-valued multi-valued (mv) logic array (see :py:mod:`~kyupy.logic`).
+            The values for primary inputs and sequential elements are filled, the primary outputs are left unassigned.
+        """
+        interface, pi_map, _, scan_maps, scan_inversions = self._maps(circuit)
+        tests = np.full((len(interface), len(self.patterns)), logic.UNASSIGNED)
         for i, p in enumerate(self.patterns):
             for si_port in self.si_ports.keys():
-                tests.set_values(i, p.load[si_port], scan_maps[si_port], scan_inversions[si_port])
-            tests.set_values(i, p.launch['_pi'], pi_map)
+                pattern = logic.mvarray(p.load[si_port][0])
+                inversions = np.choose((pattern == logic.UNASSIGNED) | (pattern == logic.UNKNOWN),
+                                       [scan_inversions[si_port], logic.ZERO]).astype(np.uint8)
+                np.bitwise_xor(pattern, inversions, out=pattern)
+                tests[scan_maps[si_port], i] = pattern
+            tests[pi_map, i] = logic.mvarray(p.capture['_pi'][0])
         return tests
 
-    def tests8v(self, c):
-        interface, pi_map, po_map, scan_maps, scan_inversions = self._maps(c)
-        init = PackedVectors(len(self.patterns), len(interface), 2)
+    def tests_loc(self, circuit, tlib, init_filter=None, launch_filter=None):
+        """Assembles and returns a LoC scan test pattern set for given circuit.
+
+        This function assumes a launch-on-capture (LoC) delay test.
+        It performs a logic simulation to obtain the first capture pattern (the one that launches the delay
+        test) and assembles the test pattern set from from pairs for initialization- and launch-patterns.
+
+        :param circuit: The circuit to assemble the patterns for. The patterns will follow the
+            :py:attr:`~kyupy.circuit.Circuit.s_nodes` ordering of the this circuit.
+        :param init_filter: A function for filtering the initialization patterns. This function is called
+            with the initialization patterns from the STIL file as mvarray before logic simulation.
+            It shall return an mvarray with the same shape. This function can be used, for example, to fill
+            patterns.
+        :param launch_filter: A function for filtering the launch patterns. This function is called
+            with the launch patterns generated by logic simulation before they are combined with
+            the initialization patterns to form the final 8-valued test patterns.
+            The function shall return an mvarray with the same shape. This function can be used, for example, to fill
+            patterns.
+        :return: An 8-valued multi-valued (mv) logic array (see :py:mod:`~kyupy.logic`). The values for primary
+            inputs and sequential elements are filled, the primary outputs are left unassigned.
+        """
+        interface, pi_map, po_map, scan_maps, scan_inversions = self._maps(circuit)
+        init = np.full((len(interface), len(self.patterns)), logic.UNASSIGNED)
         for i, p in enumerate(self.patterns):
             # init.set_values(i, '0' * len(interface))
             for si_port in self.si_ports.keys():
-                init.set_values(i, p.load[si_port], scan_maps[si_port], scan_inversions[si_port])
-            init.set_values(i, p.launch['_pi'], pi_map)
-        sim4v = LogicSim(c, len(init), 2)
-        sim4v.assign(init)
-        sim4v.propagate()
-        launch = init.copy()
-        sim4v.capture(launch)
+                pattern = logic.mvarray(p.load[si_port][0])
+                inversions = np.choose((pattern == logic.UNASSIGNED) | (pattern == logic.UNKNOWN),
+                                       [scan_inversions[si_port], logic.ZERO]).astype(np.uint8)
+                np.bitwise_xor(pattern, inversions, out=pattern)
+                init[scan_maps[si_port], i] = pattern
+            init[pi_map, i] = logic.mvarray(p.launch['_pi'][0] if '_pi' in p.launch else p.capture['_pi'][0])
+        if init_filter: init = init_filter(init)
+        circuit_resolved = circuit.copy()
+        circuit_resolved.resolve_tlib_cells(tlib)
+        sim8v = LogicSim(circuit_resolved, init.shape[-1], m=8)
+        sim8v.s[0] = logic.mv_to_bp(init)
+        sim8v.s_to_c()
+        sim8v.c_prop()
+        sim8v.c_to_s()
+        launch = logic.bp_to_mv(sim8v.s[1])[..., :init.shape[-1]]
         for i, p in enumerate(self.patterns):
-            # if there was no launch clock, then init = launch
-            if ('P' not in p.launch['_pi']) or ('P' not in p.capture['_pi']):
+            # if there was no launch cycle or launch clock, then init = launch
+            if '_pi' not in p.launch or 'P' not in p.launch['_pi'][0] or 'P' not in p.capture['_pi'][0]:
                 for si_port in self.si_ports.keys():
-                    launch.set_values(i, p.load[si_port], scan_maps[si_port], scan_inversions[si_port])
-            if '_pi' in p.capture and 'P' in p.capture['_pi']:
-                launch.set_values(i, p.capture['_pi'], pi_map)
-        
-        return PackedVectors.from_pair(init, launch)
-                
-    def responses(self, c):
-        interface, pi_map, po_map, scan_maps, scan_inversions = self._maps(c)
-        resp = PackedVectors(len(self.patterns), len(interface), 2)
+                    pattern = logic.mv_xor(logic.mvarray(p.load[si_port][0]), scan_inversions[si_port])
+                    launch[scan_maps[si_port], i] = pattern
+            if '_pi' in p.capture and 'P' in p.capture['_pi'][0]:
+                launch[pi_map, i] = logic.mvarray(p.capture['_pi'][0])
+            launch[po_map, i] = logic.UNASSIGNED
+        if launch_filter: launch = launch_filter(launch)
+
+        return logic.mv_transition(init, launch)
+
+    def responses(self, circuit):
+        """Assembles and returns a scan test response pattern set for given circuit.
+
+        :param circuit: The circuit to assemble the patterns for. The patterns will follow the
+            :py:attr:`~kyupy.circuit.Circuit.s_nodes` ordering of the this circuit.
+        :return: A 4-valued multi-valued (mv) logic array (see :py:mod:`~kyupy.logic`).
+            The values for primary outputs and sequential elements are filled, the primary inputs are left unassigned.
+        """
+        interface, _, po_map, scan_maps, scan_inversions = self._maps(circuit)
+        resp = np.full((len(interface), len(self.patterns)), logic.UNASSIGNED)
         for i, p in enumerate(self.patterns):
-            if len(p.capture) > 0:
-                resp.set_values(i, p.capture['_po'], po_map)
-            else:
-                resp.set_values(i, p.launch['_po'], po_map)
+            resp[po_map, i] = logic.mvarray(p.capture['_po'][0] if len(p.capture) > 0 else p.launch['_po'][0])
             for so_port in self.so_ports.keys():
-                resp.set_values(i, p.unload[so_port], scan_maps[so_port], scan_inversions[so_port])
+                pattern = logic.mv_xor(logic.mvarray(p.unload[so_port][0]), scan_inversions[so_port])
+                resp[scan_maps[so_port], i] = pattern
         return resp
-        
-        
+
+
 class StilTransformer(Transformer):
     def __init__(self):
         super().__init__()
         self._signal_groups = None
         self._calls = None
         self._scan_chains = None
-        
+
     @staticmethod
     def quoted(args): return args[0][1:-1]
 
     @staticmethod
     def call(args): return Call(args[0], dict(args[1:]))
-        
+
     @staticmethod
-    def call_parameter(args): return args[0], args[1].value
+    def call_parameter(args): return args[0], (args[1].value.replace('\n', '').replace('N', '-'), args[1].start_pos)
 
     @staticmethod
     def signal_group(args): return args[0], args[1:]
-    
+
     @staticmethod
     def scan_chain(args):
         scan_in = None
@@ -153,18 +213,17 @@ class StilTransformer(Transformer):
         return args[0], ([scan_in] + scan_cells + [scan_out])
 
     def signal_groups(self, args): self._signal_groups = dict(args)
-    
+
     def pattern(self, args): self._calls = [c for c in args if isinstance(c, Call)]
 
     def scan_structures(self, args): self._scan_chains = dict(args)
 
     def start(self, args):
         return StilFile(float(args[0]), self._signal_groups, self._scan_chains, self._calls)
-        
 
-def parse(stil):
-    grammar = r"""
-    start: "STIL" FLOAT _ignore _block*
+
+GRAMMAR = r"""
+    start: "STIL" FLOAT ( _ignore | ";" ) _block*
     _block: signal_groups | scan_structures | pattern
         | "Header" _ignore
         | "Signals" _ignore
@@ -173,10 +232,11 @@ def parse(stil):
         | "PatternExec" _ignore
         | "Procedures" _ignore
         | "MacroDefs" _ignore
-    
+        | "UserKeywords" /[a-zA-Z]*;/
+
     signal_groups: "SignalGroups" "{" signal_group* "}"
     signal_group: quoted "=" "'" quoted ( "+" quoted)* "'" _ignore? ";"?
-    
+
     scan_structures: "ScanStructures" "{" scan_chain* "}"
     scan_chain: "ScanChain" quoted "{" ( scan_length
         | scan_in | scan_out | scan_inversion | scan_cells | scan_master_clock )* "}"
@@ -186,7 +246,7 @@ def parse(stil):
     scan_inversion: "ScanInversion" /[0-9]+/ ";"
     scan_cells: "ScanCells" (quoted | /!/)* ";"
     scan_master_clock: "ScanMasterClock" quoted ";"
-    
+
     pattern: "Pattern" quoted "{" ( label | w | c | macro | ann | call )* "}"
     label: quoted ":"
     w: "W" quoted ";"
@@ -195,7 +255,7 @@ def parse(stil):
     ann: "Ann" _ignore
     call: "Call" quoted "{" call_parameter* "}"
     call_parameter: quoted "=" /[^;]+/ ";"
-        
+
     quoted: /"[^"]*"/
     FLOAT: /[-0-9.]+/
     _ignore: "{" _NOB? _ignore_inner* "}"
@@ -203,50 +263,16 @@ def parse(stil):
     _NOB: /[^{}]+/
     %ignore ( /\r?\n/ | "//" /[^\n]*/ | /[\t\f ]/ )+
     """
-    if '\n' not in str(stil):  # One line?: Assuming it is a file name.
-        if str(stil).endswith('.gz'):
-            with gzip.open(stil, 'rt') as f:
-                text = f.read()
-        else:
-            with open(stil, 'r') as f:
-                text = f.read()
-    else:
-        text = str(stil)
-    return Lark(grammar, parser="lalr", transformer=StilTransformer()).parse(text)
 
 
-def extract_scan_pattens(stil_calls):
-    pats = []
-    pi = None
-    scan_in = None
-    for call in stil_calls:
-        if call.name == 'load_unload':
-            scan_out = call.parameters.get('Scan_Out')
-            if scan_out is not None:
-                scan_out = scan_out.replace('\n', '')
-            if pi: pats.append(ScanPattern(scan_in, pi, None, scan_out))
-            scan_in = call.parameters.get('Scan_In')
-            if scan_in is not None:
-                scan_in = scan_in.replace('\n', '')
-        if call.name == 'allclock_capture':
-            pi = call.parameters['_pi'].replace('\n', '')
-    return pats
+def parse(text) -> StilFile:
+    """Parses the given ``text`` and returns a :class:`StilFile` object."""
+    return Lark(GRAMMAR, parser="lalr", transformer=StilTransformer()).parse(text) # type: ignore
 
 
-def match_patterns(stil_file, pats, interface):    
-    intf_pos = dict([(n.name, i) for i, n in enumerate(interface)])
-    pi_map = [intf_pos[n] for n in stil_file.signal_groups['_pi']]
-    scan_map = [intf_pos[re.sub(r'b..\.', '', n)] for n in reversed(stil_file.scan_chains['1'])]
-    # print(scan_map)
-    tests = PackedVectors(len(pats), len(interface), 2)
-    for i, p in enumerate(pats):
-        tests.set_values(i, p.scan_in, scan_map)
-        tests.set_values(i, p.pi, pi_map)
-
-    resp = PackedVectors(len(pats), len(interface), 2)
-    for i, p in enumerate(pats):
-        resp.set_values(i, p.pi, pi_map)
-        resp.set_values(i, p.scan_out, scan_map)
-
-    return tests, resp
+def load(file) -> StilFile:
+    """Parses the contents of ``file`` and returns a :class:`StilFile` object.
 
+    Files with `.gz`-suffix are decompressed on-the-fly.
+    """
+    return parse(readtext(file))

src/kyupy/techlib.py

@@ -0,0 +1,667 @@
+"""KyuPy's Built-In Technology Libraries
+
+Technology libraries provide cell definitions and their implementation with simulation primitives.
+A couple of common standard cell libraries are built-in.
+Others can be easily added by providing a bench-like description of the cells.
+"""
+
+import re
+from itertools import product
+
+from . import bench
+
+class TechLib:
+    """Class for standard cell library definitions.
+
+    :py:class:`~kyupy.circuit.Node` objects do not have pin names.
+    This class maps pin names to pin directions and defined positions in the ``node.ins`` and ``node.outs`` lists.
+    Furthermore, it gives access to implementations of complex cells. See also :py:func:`~kyupy.circuit.substitute` and
+    :py:func:`~kyupy.circuit.resolve_tlib_cells`.
+    """
+    def __init__(self, lib_src):
+        self.cells = dict()
+        """A dictionary with pin definitions and circuits for each cell kind (type).
+        """
+        for c_str in re.split(r';\s+', lib_src):
+            c_str = re.sub(r'^\s+', '', c_str)
+            name_len = c_str.find(' ')
+            if name_len <= 0: continue
+            c = bench.parse(c_str[name_len:])
+            c.name = c_str[:name_len]
+            c.eliminate_1to1_forks()
+            i_idx, o_idx = 0, 0
+            pin_dict = dict()
+            for n in c.io_nodes:
+                if len(n.ins) == 0:
+                    pin_dict[n.name] = (i_idx, False)
+                    i_idx += 1
+                else:
+                    pin_dict[n.name] = (o_idx, True)
+                    o_idx += 1
+            has_dff = 'DFF' in set(n.kind for n in c.cells.values())
+            has_latch = 'LATCH' in set(n.kind for n in c.cells.values())
+            parts = [s[1:-1].split(',') if s[0] == '{' else [s] for s in re.split(r'({[^}]+})', c.name) if len(s) > 0]
+            for name in [''.join(item) for item in product(*parts)]:
+                self.cells[name] = (c, pin_dict, has_dff, has_latch)
+
+    def pin_index(self, kind, pin):
+        """Returns a pin list position for a given node kind and pin name."""
+        assert kind in self.cells, f'Unknown cell: {kind}'
+        assert pin in self.cells[kind][1], f'Unknown pin: {pin} for cell {kind}'
+        return self.cells[kind][1][pin][0]
+
+    def pin_name(self, kind: str, pos: int, output=False) -> str:
+        """Returns the pin name for a given node kind, list position, and direction."""
+        assert kind in self.cells, f'Unknown cell: {kind}'
+        for name, (ppos, isout) in self.cells[kind][1].items():
+            if isout == output and ppos == pos:
+                return name
+        raise ValueError(f'Invalid {"output" if output else "input"} pin position {pos} for cell: {kind}')
+
+    def pin_is_output(self, kind, pin):
+        """Returns True, if given pin name of a node kind is an output."""
+        assert kind in self.cells, f'Unknown cell: {kind}'
+        assert pin in self.cells[kind][1], f'Unknown pin: {pin} for cell {kind}'
+        return self.cells[kind][1][pin][1]
+    
+    def is_dff(self, kind):
+        """Returns True, if given node kind is a d-flip-flop."""
+        if kind == '__fork__': return False
+        if kind == '__const0__': return False
+        if kind == '__const1__': return False
+        if kind == 'input': return False
+        if kind == 'output': return False
+        assert kind in self.cells, f'Unknown cell: {kind}'
+        return self.cells[kind][2]
+    
+    def is_latch(self, kind):
+        """Returns True, if given node kind is a latch."""
+        if kind == '__fork__': return False
+        if kind == '__const0__': return False
+        if kind == '__const1__': return False
+        if kind == 'input': return False
+        if kind == 'output': return False
+        assert kind in self.cells, f'Unknown cell: {kind}'
+        return self.cells[kind][3]
+
+
+KYUPY = TechLib(r"""
+BUF1    input(i0)          output(o) o=BUF1(i0)          ;
+INV1    input(i0)          output(o) o=INV1(i0)          ;
+AND2    input(i0,i1)       output(o) o=AND2(i0,i1)       ;
+AND3    input(i0,i1,i2)    output(o) o=AND3(i0,i1,i2)    ;
+AND4    input(i0,i1,i2,i3) output(o) o=AND4(i0,i1,i2,i3) ;
+NAND2   input(i0,i1)       output(o) o=NAND2(i0,i1)      ;
+NAND3   input(i0,i1,i2)    output(o) o=NAND3(i0,i1,i2)   ;
+NAND4   input(i0,i1,i2,i3) output(o) o=NAND4(i0,i1,i2,i3);
+OR2     input(i0,i1)       output(o) o=OR2(i0,i1)        ;
+OR3     input(i0,i1,i2)    output(o) o=OR3(i0,i1,i2)     ;
+OR4     input(i0,i1,i2,i3) output(o) o=OR4(i0,i1,i2,i3)  ;
+NOR2    input(i0,i1)       output(o) o=NOR2(i0,i1)       ;
+NOR3    input(i0,i1,i2)    output(o) o=NOR3(i0,i1,i2)    ;
+NOR4    input(i0,i1,i2,i3) output(o) o=NOR4(i0,i1,i2,i3) ;
+XOR2    input(i0,i1)       output(o) o=XOR2(i0,i1)       ;
+XOR3    input(i0,i1,i2)    output(o) o=XOR3(i0,i1,i2)    ;
+XOR4    input(i0,i1,i2,i3) output(o) o=XOR4(i0,i1,i2,i3) ;
+XNOR2   input(i0,i1)       output(o) o=XNOR2(i0,i1)      ;
+XNOR3   input(i0,i1,i2)    output(o) o=XNOR3(i0,i1,i2)   ;
+XNOR4   input(i0,i1,i2,i3) output(o) o=XNOR4(i0,i1,i2,i3)  ;
+AO21    input(i0,i1,i2)    output(o) o=AO21(i0,i1,i2)      ;
+AO22    input(i0,i1,i2,i3) output(o) o=AO22(i0,i1,i2,i3)   ;
+OA21    input(i0,i1,i2)    output(o) o=OA21(i0,i1,i2)      ;
+OA22    input(i0,i1,i2,i3) output(o) o=OA22(i0,i1,i2,i3)   ;
+AOI21   input(i0,i1,i2)    output(o) o=AOI21(i0,i1,i2)     ;
+AOI22   input(i0,i1,i2,i3) output(o) o=AOI22(i0,i1,i2,i3)  ;
+OAI21   input(i0,i1,i2)    output(o) o=OAI21(i0,i1,i2)     ;
+OAI22   input(i0,i1,i2,i3) output(o) o=OAI22(i0,i1,i2,i3)  ;
+AO211   input(i0,i1,i2,i3) output(o) o=AO211(i0,i1,i2,i3)  ;
+OA211   input(i0,i1,i2,i3) output(o) o=OA211(i0,i1,i2,i3)  ;
+AOI211  input(i0,i1,i2,i3) output(o) o=AOI211(i0,i1,i2,i3) ;
+OAI211  input(i0,i1,i2,i3) output(o) o=OAI211(i0,i1,i2,i3) ;
+MUX21   input(i0,i1,i2)    output(o) o=MUX21(i0,i1,i2)     ;
+DFF     input(D,CLK)       output(Q) Q=DFF(D,CLK)        ;
+LATCH   input(D,CLK)       output(Q) Q=LATCH(D,CLK)      ;
+""")
+"""A synthetic library of all KyuPy simulation primitives.
+"""
+
+GSC180 = TechLib(r"""
+BUFX{1,3}      input(A)    output(Y) Y=BUF1(A)    ;
+CLKBUFX{1,2,3} input(A)    output(Y) Y=BUF1(A)    ;
+INVX{1,2,4,8}  input(A)    output(Y) Y=INV1(A)    ;
+TBUFX{1,2,4,8} input(A,OE) output(Y) Y=AND2(A,OE) ;
+TINVX1         input(A,OE) output(Y) AB=INV1(A) Y=AND2(AB,OE) ;
+
+AND2X1      input(A,B)     output(Y) Y=AND2(A,B)      ;
+NAND2X{1,2} input(A,B)     output(Y) Y=NAND2(A,B)     ;
+NAND3X1     input(A,B,C)   output(Y) Y=NAND3(A,B,C)   ;
+NAND4X1     input(A,B,C,D) output(Y) Y=NAND4(A,B,C,D) ;
+OR2X1       input(A,B)     output(Y) Y=OR2(A,B)       ;
+OR4X1       input(A,B,C,D) output(Y) Y=OR4(A,B,C,D)   ;
+NOR2X1      input(A,B)     output(Y) Y=NOR2(A,B)      ;
+NOR3X1      input(A,B,C)   output(Y) Y=NOR3(A,B,C)    ;
+NOR4X1      input(A,B,C,D) output(Y) Y=NOR4(A,B,C,D)  ;
+XOR2X1      input(A,B)     output(Y) Y=XOR2(A,B)      ;
+
+MX2X1   input(A,B,S0)            output(Y)    Y=MUX21(A,B,S0)      ;
+AOI21X1 input(A0,A1,B0)          output(Y)    Y=AOI21(A0,A1,B0)    ;
+AOI22X1 input(A0,A1,B0,B1)       output(Y)    Y=AOI22(A0,A1,B0,B1) ;
+OAI21X1 input(A0,A1,B0)          output(Y)    Y=OAI21(A0,A1,B0)    ;
+OAI22X1 input(A0,A1,B0,B1)       output(Y)    Y=OAI22(A0,A1,B0,B1) ;
+OAI33X1 input(A0,A1,A2,B0,B1,B2) output(Y)    AA=OR2(A0,A1) BB=OR2(B0,B1) Y=OAI22(AA,A2,BB,B2) ;
+ADDFX1  input(A,B,CI)            output(CO,S) AB=XOR2(A,B) CO=XOR2(AB,CI) S=AO22(AB,CI,A,B)    ;
+ADDHX1  input(A,B)               output(CO,S) CO=AND2(A,B) S=XOR2(A,B)                         ;
+
+DFFX1    input(CK,D)             output(Q,QN) Q=DFF(D,CK) QN=INV1(Q) ;
+DFFSRX1  input(CK,D,RN,SN)       output(Q,QN) DR=AND2(D,RN) SET=INV1(SN) DRS=OR2(DR,SET) Q=DFF(DRS,CK) QN=INV1(Q) ;
+SDFFSRX1 input(CK,D,RN,SE,SI,SN) output(Q,QN) DR=AND2(D,RN) SET=INV1(SN) DRS=OR2(DR,SET) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CK) QN=INV1(Q) ;
+
+TLATSRX1 input(D,G,RN,SN) output(Q,QN) DR=AND2(D,RN) SET=INV1(SN) DRS=OR2(DR,SET) Q=LATCH(DRS,G) QN=INV1(Q) ;
+TLATX1   input(C,D)       output(Q,QN) Q=LATCH(D,C) QN=INV1(Q) ;
+""")
+"""The GSC 180nm generic standard cell library.
+"""
+
+
+NANGATE = TechLib(r"""
+FILLTIE ;
+FILL_X{1,2,4,8,16} ;
+ANTENNA input(I)   ;
+
+TIEH output(Z) Z=__const1__() ;
+TIEL output(ZN) ZN=__const0__() ;
+
+BUF_X{1,2,4,8,12,16}  input(I) output(Z)  Z=BUF1(I)  ;
+INV_X{1,2,4,8,12,16}  input(I) output(ZN) ZN=INV1(I) ;
+
+CLKBUF_X{1,2,4,8,12,16} input(I) output(Z)  Z=BUF1(I)  ;
+CLKGATETST_X1 input(CLK,E,TE) output(Q) Q=OA21(CLK,E,TE) ;
+
+AND2_X{1,2}  input(A1,A2)       output(Z)  Z=AND2(A1,A2)         ;
+AND3_X{1,2}  input(A1,A2,A3)    output(Z)  Z=AND3(A1,A2,A3)      ;
+AND4_X{1,2}  input(A1,A2,A3,A4) output(Z)  Z=AND4(A1,A2,A3,A4)   ;
+NAND2_X{1,2} input(A1,A2)       output(ZN) ZN=NAND2(A1,A2)       ;
+NAND3_X{1,2} input(A1,A2,A3)    output(ZN) ZN=NAND3(A1,A2,A3)    ;
+NAND4_X{1,2} input(A1,A2,A3,A4) output(ZN) ZN=NAND4(A1,A2,A3,A4) ;
+OR2_X{1,2}   input(A1,A2)       output(Z)  Z=OR2(A1,A2)          ;
+OR3_X{1,2}   input(A1,A2,A3)    output(Z)  Z=OR3(A1,A2,A3)       ;
+OR4_X{1,2}   input(A1,A2,A3,A4) output(Z)  Z=OR4(A1,A2,A3,A4)    ;
+NOR2_X{1,2}  input(A1,A2)       output(ZN) ZN=NOR2(A1,A2)        ;
+NOR3_X{1,2}  input(A1,A2,A3)    output(ZN) ZN=NOR3(A1,A2,A3)     ;
+NOR4_X{1,2}  input(A1,A2,A3,A4) output(ZN) ZN=NOR4(A1,A2,A3,A4)  ;
+XOR2_X1      input(A1,A2)       output(Z)  Z=XOR2(A1,A2)         ;
+XNOR2_X1     input(A1,A2)       output(ZN) ZN=XNOR2(A1,A2)       ;
+
+MUX2_X1 input(I0,I1,S) output(Z) Z=MUX21(I0,I1,S) ;
+
+HA_X1 input(A,B) output(CO,S) CO=AND2(A,B) S=XOR2(A,B) ;
+FA_X1 input(A,B,CI) output(CO,S) AB=XOR2(A,B) CO=XOR2(AB,CI) S=AO22(CI,A,B) ;
+
+AOI21_X{1,2} input(A1,A2,B)     output(ZN) ZN=AOI21(A1,A2,B)     ;
+OAI21_X{1,2} input(A1,A2,B)     output(ZN) ZN=OAI21(A1,A2,B)     ;
+AOI22_X{1,2} input(A1,A2,B1,B2) output(ZN) ZN=AOI22(A1,A2,B1,B2) ;
+OAI22_X{1,2} input(A1,A2,B1,B2) output(ZN) ZN=OAI22(A1,A2,B1,B2) ;
+
+DFFRNQ_X1  input(D,RN,CLK)    output(Q)  DR=AND2(D,RN) Q=DFF(DR,CLK) ;
+DFFSNQ_X1  input(D,SN,CLK)    output(Q)  S=INV1(SN) DS=OR2(D,S) Q=DFF(DS,CLK) ;
+
+SDFFRNQ_X1 input(D,RN,SE,SI,CLK) output(Q) DR=AND2(D,RN) DI=MUX21(DR,SI,SE) Q=DFF(DI,CLK) ;
+SDFFSNQ_X1 input(D,SE,SI,SN,CLK) output(Q) S=INV1(SN) DS=OR2(D,S) DI=MUX21(DS,SI,SE) Q=DFF(DI,CLK) ;
+
+TBUF_X{1,2,4,8,12,16} input(EN,I)   output(Z)  Z=BUF1(I) ;
+
+LHQ_X1 input(D,E) output(Q)  Q=LATCH(D,E) ;
+""")
+"""Nangate 15nm Open Cell Library (NanGate_15nm_OCL_v0.1_2014_06.A).
+"""
+
+
+NANGATE45 = TechLib(r"""
+FILLCELL_X{1,2,4,8,16,32} ;
+ANTENNA_X1 input(A) ;
+
+LOGIC0_X1 output(Z) Z=__const0__() ;
+LOGIC1_X1 output(Z) Z=__const1__() ;
+
+BUF_X{1,2,4,8,16,32}  input(A) output(Z)  Z=BUF1(A)  ;
+INV_X{1,2,4,8,16,32}  input(A) output(ZN) ZN=INV1(A) ;
+
+CLKBUF_X{1,2,3}       input(A)       output(Z)   Z=BUF1(A)         ;
+CLKGATETST_X{1,2,4,8} input(CK,E,SE) output(GCK) GCK=OA21(CK,E,SE) ;
+CLKGATE_X{1,2,4,8}    input(CK,E)    output(GCK) GCK=AND2(CK,E)    ;
+
+AND2_X{1,2,4}  input(A1,A2)       output(ZN) ZN=AND2(A1,A2)        ;
+AND3_X{1,2,4}  input(A1,A2,A3)    output(ZN) ZN=AND3(A1,A2,A3)     ;
+AND4_X{1,2,4}  input(A1,A2,A3,A4) output(ZN) ZN=AND4(A1,A2,A3,A4)  ;
+NAND2_X{1,2,4} input(A1,A2)       output(ZN) ZN=NAND2(A1,A2)       ;
+NAND3_X{1,2,4} input(A1,A2,A3)    output(ZN) ZN=NAND3(A1,A2,A3)    ;
+NAND4_X{1,2,4} input(A1,A2,A3,A4) output(ZN) ZN=NAND4(A1,A2,A3,A4) ;
+OR2_X{1,2,4}   input(A1,A2)       output(ZN) ZN=OR2(A1,A2)         ;
+OR3_X{1,2,4}   input(A1,A2,A3)    output(ZN) ZN=OR3(A1,A2,A3)      ;
+OR4_X{1,2,4}   input(A1,A2,A3,A4) output(ZN) ZN=OR4(A1,A2,A3,A4)   ;
+NOR2_X{1,2,4}  input(A1,A2)       output(ZN) ZN=NOR2(A1,A2)        ;
+NOR3_X{1,2,4}  input(A1,A2,A3)    output(ZN) ZN=NOR3(A1,A2,A3)     ;
+NOR4_X{1,2,4}  input(A1,A2,A3,A4) output(ZN) ZN=NOR4(A1,A2,A3,A4)  ;
+XOR2_X{1,2}    input(A,B)         output(Z)  Z=XOR2(A,B)           ;
+XNOR2_X{1,2}   input(A,B)         output(ZN) ZN=XNOR2(A,B)         ;
+
+MUX2_X{1,2} input(A,B,S) output(Z) Z=MUX21(A,B,S) ;
+
+HA_X1 input(A,B) output(CO,S) CO=AND2(A,B) S=XOR2(A,B) ;
+FA_X1 input(A,B,CI) output(CO,S) AB=XOR2(A,B) CO=XOR2(AB,CI) S=AO22(CI,A,B) ;
+
+AOI21_X{1,2,4} input(A,B1,B2)     output(ZN) ZN=AOI21(B1,B2,A)     ;
+OAI21_X{1,2,4} input(A,B1,B2)     output(ZN) ZN=OAI21(B1,B2,A)     ;
+AOI22_X{1,2,4} input(A1,A2,B1,B2) output(ZN) ZN=AOI22(A1,A2,B1,B2) ;
+OAI22_X{1,2,4} input(A1,A2,B1,B2) output(ZN) ZN=OAI22(A1,A2,B1,B2) ;
+
+OAI211_X{1,2,4} input(A,B,C1,C2) output(ZN) ZN=OAI211(C1,C2,A,B)   ;
+AOI211_X{1,2,4} input(A,B,C1,C2) output(ZN) ZN=AOI211(C1,C2,A,B)   ;
+
+AOI221_X{1,2,4} input(A,B1,B2,C1,C2) output(ZN) BC=AO22(B1,B2,C1,C2) ZN=NOR2(BC,A)  ;
+OAI221_X{1,2,4} input(A,B1,B2,C1,C2) output(ZN) BC=OA22(B1,B2,C1,C2) ZN=NAND2(BC,A) ;
+
+AOI222_X{1,2,4} input(A1,A2,B1,B2,C1,C2) output(ZN) BC=AO22(B1,B2,C1,C2) ZN=AOI21(A1,A2,BC) ;
+OAI222_X{1,2,4} input(A1,A2,B1,B2,C1,C2) output(ZN) BC=OA22(B1,B2,C1,C2) ZN=OAI21(A1,A2,BC) ;
+
+OAI33_X1 input(A1,A2,A3,B1,B2,B3) output(ZN) AA=OR2(A1,A2) BB=OR2(B1,B2) ZN=OAI22(AA,A3,BB,B3) ;
+
+DFF_X{1,2}   input(D,CK)       output(Q,QN)  Q=DFF(D,CK) QN=INV1(Q) ;
+DFFR_X{1,2}  input(D,RN,CK)    output(Q,QN)  DR=AND2(D,RN) Q=DFF(DR,CK) QN=INV1(Q) ;
+DFFS_X{1,2}  input(D,SN,CK)    output(Q,QN)  S=INV1(SN) DS=OR2(D,S) Q=DFF(DS,CK) QN=INV1(Q) ;
+DFFRS_X{1,2} input(D,RN,SN,CK) output(Q,QN)  S=INV1(SN) DS=OR2(D,S) DRS=AND2(DS,RN) Q=DFF(DRS,CK) QN=INV1(Q) ;
+
+SDFF_X{1,2}   input(D,SE,SI,CK)       output(Q,QN)  DI=MUX21(D,SI,SE) Q=DFF(DI,CK) QN=INV1(Q) ;
+SDFFR_X{1,2}  input(D,RN,SE,SI,CK)    output(Q,QN)  DR=AND2(D,RN) DI=MUX21(DR,SI,SE) Q=DFF(DI,CK) QN=INV1(Q) ;
+SDFFS_X{1,2}  input(D,SE,SI,SN,CK)    output(Q,QN)  S=INV1(SN) DS=OR2(D,S) DI=MUX21(DS,SI,SE) Q=DFF(DI,CK) QN=INV1(Q) ;
+SDFFRS_X{1,2} input(D,RN,SE,SI,SN,CK) output(Q,QN)  S=INV1(SN) DS=OR2(D,S) DRS=AND2(DS,RN) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CK) QN=INV1(Q) ;
+
+TBUF_X{1,2,4,8,16} input(A,EN)   output(Z)  Z=BUF1(A)    ;
+
+TINV_X1            input(I,EN)   output(ZN) ZN=INV1(I)   ;
+TLAT_X1            input(D,G,OE) output(Q)  Q=LATCH(D,G) ;
+
+DLH_X{1,2} input(D,G) output(Q)  Q=LATCH(D,G)            ;
+DLL_X{1,2} input(D,GN) output(Q) G=INV1(GN) Q=LATCH(D,G) ;
+
+""")
+"""Nangate 45nm Open Cell Library (NangateOpenCellLibrary_PDKv1_3_v2010_12).
+This NANGATE-variant that uses 'ZN' as output pin names for AND and OR gates.
+"""
+
+
+SAED32 = TechLib(r"""
+NBUFFX{2,4,8,16,32}$ input(A) output(Y) Y=BUF1(A) ;
+AOBUFX{1,2,4}$       input(A) output(Y) Y=BUF1(A) ;
+DELLN{1,2,3}X2$      input(A) output(Y) Y=BUF1(A) ;
+
+INVX{0,1,2,4,8,16,32}$ input(A) output(Y) Y=INV1(A) ;
+AOINVX{1,2,4}$         input(A) output(Y) Y=INV1(A) ;
+IBUFFX{2,4,8,16,32}$   input(A) output(Y) Y=INV1(A) ;
+
+TIEH$ output(Y) Y=__const1__() ;
+TIEL$ output(Y) Y=__const0__() ;
+
+HEAD2X{2,4,8,16,32}$ input(SLEEP) output(SLEEPOUT) SLEEPOUT=BUF1(SLEEP) ;
+HEADX{2,4,8,16,32}$  input(SLEEP) ;
+
+FOOT2X{2,4,8,16,32}$ input(SLEEP) output(SLEEPOUT) SLEEPOUT=BUF1(SLEEP) ;
+FOOTX{2,4,8,16,32}$  input(SLEEP) ;
+
+ANTENNA$ input(INP)   ;
+CLOAD1$  input(A)     ;
+DCAP$                 ;
+DHFILLH2$             ;
+DHFILLHL2$            ;
+DHFILLHLHLS11$        ;
+SHFILL{1,2,3,64,128}$ ;
+
+AND2X{1,2,4}$    input(A1,A2)       output(Y) Y=AND2(A1,A2)        ;
+AND3X{1,2,4}$    input(A1,A2,A3)    output(Y) Y=AND3(A1,A2,A3)     ;
+AND4X{1,2,4}$    input(A1,A2,A3,A4) output(Y) Y=AND4(A1,A2,A3,A4)  ;
+OR2X{1,2,4}$     input(A1,A2)       output(Y) Y=OR2(A1,A2)         ;
+OR3X{1,2,4}$     input(A1,A2,A3)    output(Y) Y=OR3(A1,A2,A3)      ;
+OR4X{1,2,4}$     input(A1,A2,A3,A4) output(Y) Y=OR4(A1,A2,A3,A4)   ;
+XOR2X{1,2}$      input(A1,A2)       output(Y) Y=XOR2(A1,A2)        ;
+XOR3X{1,2}$      input(A1,A2,A3)    output(Y) Y=XOR3(A1,A2,A3)     ;
+NAND2X{0,1,2,4}$ input(A1,A2)       output(Y) Y=NAND2(A1,A2)       ;
+NAND3X{0,1,2,4}$ input(A1,A2,A3)    output(Y) Y=NAND3(A1,A2,A3)    ;
+NAND4X{0,1}$     input(A1,A2,A3,A4) output(Y) Y=NAND4(A1,A2,A3,A4) ;
+NOR2X{0,1,2,4}$  input(A1,A2)       output(Y) Y=NOR2(A1,A2)        ;
+NOR3X{0,1,2,4}$  input(A1,A2,A3)    output(Y) Y=NOR3(A1,A2,A3)     ;
+NOR4X{0,1}$      input(A1,A2,A3,A4) output(Y) Y=NOR4(A1,A2,A3,A4)  ;
+XNOR2X{1,2}$     input(A1,A2)       output(Y) Y=XNOR2(A1,A2)       ;
+XNOR3X{1,2}$     input(A1,A2,A3)    output(Y) Y=XNOR3(A1,A2,A3)    ;
+
+ISOLAND{,AO}X{1,2,4,8}$ input(ISO,D) output(Q) ISOB=NOT1(ISO) Q=AND2(ISOB,D) ;
+ISOLOR{,AO}X{1,2,4,8}$  input(ISO,D) output(Q) Q=OR2(ISO,D)  ;
+
+AO21X{1,2}$  input(A1,A2,A3) output(Y) Y=AO21(A1,A2,A3)  ;
+OA21X{1,2}$  input(A1,A2,A3) output(Y) Y=OA21(A1,A2,A3)  ;
+AOI21X{1,2}$ input(A1,A2,A3) output(Y) Y=AOI21(A1,A2,A3) ;
+OAI21X{1,2}$ input(A1,A2,A3) output(Y) Y=OAI21(A1,A2,A3) ;
+
+AO22X{1,2}$  input(A1,A2,A3,A4) output(Y) Y=AO22(A1,A2,A3,A4)  ;
+OA22X{1,2}$  input(A1,A2,A3,A4) output(Y) Y=OA22(A1,A2,A3,A4)  ;
+AOI22X{1,2}$ input(A1,A2,A3,A4) output(Y) Y=AOI22(A1,A2,A3,A4) ;
+OAI22X{1,2}$ input(A1,A2,A3,A4) output(Y) Y=OAI22(A1,A2,A3,A4) ;
+
+MUX21X{1,2}$ input(A1,A2,S0) output(Y) Y=MUX21(A1,A2,S0) ;
+
+AO221X{1,2}$  input(A1,A2,A3,A4,A5) output(Y) A=AO22(A1,A2,A3,A4) Y=OR2(A5,A)   ;
+OA221X{1,2}$  input(A1,A2,A3,A4,A5) output(Y) A=OA22(A1,A2,A3,A4) Y=AND2(A5,A)  ;
+AOI221X{1,2}$ input(A1,A2,A3,A4,A5) output(Y) A=AO22(A1,A2,A3,A4) Y=NOR2(A5,A)  ;
+OAI221X{1,2}$ input(A1,A2,A3,A4,A5) output(Y) A=OA22(A1,A2,A3,A4) Y=NAND2(A5,A) ;
+
+AO222X{1,2}$ input(A1,A2,A3,A4,A5,A6)  output(Y) A=AO22(A1,A2,A3,A4) Y=AO21(A5,A6,A)  ;
+OA222X{1,2}$ input(A1,A2,A3,A4,A5,A6)  output(Y) A=OA22(A1,A2,A3,A4) Y=OA21(A5,A6,A)  ;
+AOI222X{1,2}$ input(A1,A2,A3,A4,A5,A6) output(Y) A=AO22(A1,A2,A3,A4) Y=AOI21(A5,A6,A) ;
+OAI222X{1,2}$ input(A1,A2,A3,A4,A5,A6) output(Y) A=OA22(A1,A2,A3,A4) Y=OAI21(A5,A6,A) ;
+
+MUX41X{1,2}$ input(A1,A2,A3,A4,S0,S1) output(Y) A=MUX21(A1,A2,S0) B=MUX21(A3,A4,S0) Y=MUX21(A,B,S1) ;
+
+DEC24X{1,2}$ input(A0,A1) output(Y0,Y1,Y2,Y3) A0B=INV1(A0) A1B=INV1(A1) Y0=NOR2(A0,A1) Y1=AND(A0,A1B) Y2=AND(A0B,A1) Y3=AND(A0,A1) ;
+FADDX{1,2}$ input(A,B,CI) output(S,CO) AB=XOR2(A,B) CO=XOR2(AB,CI) S=AO22(AB,CI,A,B) ;
+HADDX{1,2}$ input(A0,B0) output(SO,C1) C1=AND2(A0,B0) SO=XOR2(A0,B0) ;
+
+{,AO}DFFARX{1,2}$ input(D,CLK,RSTB)      output(Q,QN) DR=AND2(D,RSTB) Q=DFF(DR,CLK) QN=INV1(Q) ;
+DFFASRX{1,2}$     input(D,CLK,RSTB,SETB) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) Q=DFF(DRS,CLK) QN=INV1(Q) ;
+DFFASX{1,2}$      input(D,CLK,SETB)      output(Q,QN) SET=INV1(SETB) DS=OR2(D,SET) Q=DFF(DS,CLK) QN=INV1(Q) ;
+DFFSSRX{1,2}$     input(CLK,D,RSTB,SETB) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) Q=DFF(DRS,CLK) QN=INV1(Q) ;
+DFFX{1,2}$        input(D,CLK)           output(Q,QN) Q=DFF(D,CLK) QN=INV1(Q) ;
+
+SDFFARX{1,2}$   input(D,CLK,RSTB,SE,SI)      output(Q,QN) DR=AND2(D,RSTB) DI=MUX21(DR,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFASRSX{1,2}$ input(D,CLK,RSTB,SETB,SE,SI) output(Q,QN,SO) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) SO=BUF1(Q) ;
+SDFFASRX{1,2}$  input(D,CLK,RSTB,SETB,SE,SI) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFASX{1,2}$   input(D,CLK,SETB,SE,SI)      output(Q,QN) SET=INV1(SETB) DS=OR2(D,SET) DI=MUX21(DS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFSSRX{1,2}$  input(CLK,D,RSTB,SETB,SI,SE) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFX{1,2}$     input(D,CLK,SE,SI)           output(Q,QN) DI=MUX21(D,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+
+LATCHX{1,2}$ input(D,CLK) output(Q,QN) Q=LATCH(D,CLK) QN=INV1(Q) ;
+""".replace('$','_RVT'))
+"""The SAED 32nm educational technology library.
+It defines all cells except: negative-edge flip-flops, tri-state, latches, clock gating, level shifters
+"""
+
+
+SAED90 = TechLib(r"""
+NBUFFX{2,4,8,16,32}$ input(INP) output(Z) Z=BUF1(INP) ;
+AOBUFX{1,2,4}$       input(INP) output(Z) Z=BUF1(INP) ;
+DELLN{1,2,3}X2$      input(INP) output(Z) Z=BUF1(INP) ;
+
+INVX{0,1,2,4,8,16,32}$ input(INP) output(ZN) ZN=INV1(INP) ;
+AOINVX{1,2,4}$         input(INP) output(ZN) ZN=INV1(INP) ;
+IBUFFX{2,4,8,16,32}$   input(INP) output(ZN) ZN=INV1(INP) ;
+
+TIEH$ output(Z)   Z=__const1__() ;
+TIEL$ output(ZN) ZN=__const0__() ;
+
+HEAD2X{2,4,8,16,32}$ input(SLEEP) output(SLEEPOUT) SLEEPOUT=BUF1(SLEEP) ;
+HEADX{2,4,8,16,32}$  input(SLEEP) ;
+
+ANTENNA$ input(INP)   ;
+CLOAD1$  input(INP)   ;
+DCAP$                 ;
+DHFILL{HLH,LHL}2      ;
+DHFILLHLHLS11$        ;
+SHFILL{1,2,3,64,128}$ ;
+
+AND2X{1,2,4}$    input(IN1,IN2)         output(Q)   Q=AND2(IN1,IN2)          ;
+AND3X{1,2,4}$    input(IN1,IN2,IN3)     output(Q)   Q=AND3(IN1,IN2,IN3)      ;
+AND4X{1,2,4}$    input(IN1,IN2,IN3,IN4) output(Q)   Q=AND4(IN1,IN2,IN3,IN4)  ;
+OR2X{1,2,4}$     input(IN1,IN2)         output(Q)   Q=OR2(IN1,IN2)           ;
+OR3X{1,2,4}$     input(IN1,IN2,IN3)     output(Q)   Q=OR3(IN1,IN2,IN3)       ;
+OR4X{1,2,4}$     input(IN1,IN2,IN3,IN4) output(Q)   Q=OR4(IN1,IN2,IN3,IN4)   ;
+XOR2X{1,2}$      input(IN1,IN2)         output(Q)   Q=XOR2(IN1,IN2)          ;
+XOR3X{1,2}$      input(IN1,IN2,IN3)     output(Q)   Q=XOR3(IN1,IN2,IN3)      ;
+NAND2X{0,1,2,4}$ input(IN1,IN2)         output(QN) QN=NAND2(IN1,IN2)         ;
+NAND3X{0,1,2,4}$ input(IN1,IN2,IN3)     output(QN) QN=NAND3(IN1,IN2,IN3)     ;
+NAND4X{0,1}$     input(IN1,IN2,IN3,IN4) output(QN) QN=NAND4(IN1,IN2,IN3,IN4) ;
+NOR2X{0,1,2,4}$  input(IN1,IN2)         output(QN) QN=NOR2(IN1,IN2)          ;
+NOR3X{0,1,2,4}$  input(IN1,IN2,IN3)     output(QN) QN=NOR3(IN1,IN2,IN3)      ;
+NOR4X{0,1}$      input(IN1,IN2,IN3,IN4) output(QN) QN=NOR4(IN1,IN2,IN3,IN4)  ;
+XNOR2X{1,2}$     input(IN1,IN2)         output(Q)   Q=XNOR2(IN1,IN2)         ;
+XNOR3X{1,2}$     input(IN1,IN2,IN3)     output(Q)   Q=XNOR3(IN1,IN2,IN3)     ;
+
+ISOLAND{,AO}X{1,2,4,8}$ input(ISO,D) output(Q) ISOB=NOT1(ISO) Q=AND2(ISOB,D) ;
+ISOLOR{,AO}X{1,2,4,8}$  input(ISO,D) output(Q) Q=OR2(ISO,D)  ;
+
+AO21X{1,2}$  input(IN1,IN2,IN3) output(Q)   Q=AO21(IN1,IN2,IN3)  ;
+OA21X{1,2}$  input(IN1,IN2,IN3) output(Q)   Q=OA21(IN1,IN2,IN3)  ;
+AOI21X{1,2}$ input(IN1,IN2,IN3) output(QN) QN=AOI21(IN1,IN2,IN3) ;
+OAI21X{1,2}$ input(IN1,IN2,IN3) output(QN) QN=OAI21(IN1,IN2,IN3) ;
+
+AO22X{1,2}$  input(IN1,IN2,IN3,IN4) output(Q)   Q=AO22(IN1,IN2,IN3,IN4)  ;
+OA22X{1,2}$  input(IN1,IN2,IN3,IN4) output(Q)   Q=OA22(IN1,IN2,IN3,IN4)  ;
+AOI22X{1,2}$ input(IN1,IN2,IN3,IN4) output(QN) QN=AOI22(IN1,IN2,IN3,IN4) ;
+OAI22X{1,2}$ input(IN1,IN2,IN3,IN4) output(QN) QN=OAI22(IN1,IN2,IN3,IN4) ;
+
+MUX21X{1,2}$ input(IN1,IN2,S) output(Q) Q=MUX21(IN1,IN2,S) ;
+
+AO221X{1,2}$  input(IN1,IN2,IN3,IN4,IN5) output(Q)  A=AO22(IN1,IN2,IN3,IN4)  Q=OR2(IN5,A)   ;
+OA221X{1,2}$  input(IN1,IN2,IN3,IN4,IN5) output(Q)  A=OA22(IN1,IN2,IN3,IN4)  Q=AND2(IN5,A)  ;
+AOI221X{1,2}$ input(IN1,IN2,IN3,IN4,IN5) output(QN) A=AO22(IN1,IN2,IN3,IN4) QN=NOR2(IN5,A)  ;
+OAI221X{1,2}$ input(IN1,IN2,IN3,IN4,IN5) output(QN) A=OA22(IN1,IN2,IN3,IN4) QN=NAND2(IN5,A) ;
+
+AO222X{1,2}$ input(IN1,IN2,IN3,IN4,IN5,IN6)  output(Q)  A=AO22(IN1,IN2,IN3,IN4)  Q=AO21(IN5,IN6,A)  ;
+OA222X{1,2}$ input(IN1,IN2,IN3,IN4,IN5,IN6)  output(Q)  A=OA22(IN1,IN2,IN3,IN4)  Q=OA21(IN5,IN6,A)  ;
+AOI222X{1,2}$ input(IN1,IN2,IN3,IN4,IN5,IN6) output(QN) A=AO22(IN1,IN2,IN3,IN4) QN=AOI21(IN5,IN6,A) ;
+OAI222X{1,2}$ input(IN1,IN2,IN3,IN4,IN5,IN6) output(QN) A=OA22(IN1,IN2,IN3,IN4) QN=OAI21(IN5,IN6,A) ;
+
+MUX41X{1,2}$ input(IN1,IN2,IN3,IN4,S0,S1) output(Q) A=MUX21(IN1,IN2,S0) B=MUX21(IN3,IN4,S0) Q=MUX21(A,B,S1) ;
+
+DEC24X{1,2}$ input(IN1,IN2) output(Q0,Q1,Q2,Q3) IN1B=INV1(IN1) IN2B=INV1(IN2) Q0=NOR2(IN1,IN2) Q1=AND(IN1,IN2B) Q2=AND(IN1B,IN2) Q3=AND(IN1,IN2) ;
+FADDX{1,2}$ input(A,B,CI) output(S,CO) AB=XOR2(A,B) CO=XOR2(AB,CI) S=AO22(AB,CI,A,B) ;
+HADDX{1,2}$ input(A0,B0) output(SO,C1) C1=AND2(A0,B0) SO=XOR2(A0,B0) ;
+
+{,AO}DFFARX{1,2}$ input(D,CLK,RSTB)      output(Q,QN) DR=AND2(D,RSTB) Q=DFF(DR,CLK) QN=INV1(Q) ;
+DFFASRX{1,2}$     input(D,CLK,RSTB,SETB) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) Q=DFF(DRS,CLK) QN=INV1(Q) ;
+DFFASX{1,2}$      input(D,CLK,SETB)      output(Q,QN) SET=INV1(SETB) DS=OR2(D,SET) Q=DFF(DS,CLK) QN=INV1(Q) ;
+DFFSSRX{1,2}$     input(CLK,D,RSTB,SETB) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) Q=DFF(DRS,CLK) QN=INV1(Q) ;
+DFFX{1,2}$        input(D,CLK)           output(Q,QN) Q=DFF(D,CLK) QN=INV1(Q) ;
+
+SDFFARX{1,2}$   input(D,CLK,RSTB,SE,SI)      output(Q,QN) DR=AND2(D,RSTB) DI=MUX21(DR,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFASRSX{1,2}$ input(D,CLK,RSTB,SETB,SE,SI) output(Q,QN,S0) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) S0=BUF1(Q) ;
+SDFFASRX{1,2}$  input(D,CLK,RSTB,SETB,SE,SI) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFASX{1,2}$   input(D,CLK,SETB,SE,SI)      output(Q,QN) SET=INV1(SETB) DS=OR2(D,SET) DI=MUX21(DS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFSSRX{1,2}$  input(CLK,D,RSTB,SETB,SI,SE) output(Q,QN) DR=AND2(D,RSTB) SET=INV1(SETB) DRS=OR2(DR,SET) DI=MUX21(DRS,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+SDFFX{1,2}$     input(D,CLK,SE,SI)           output(Q,QN) DI=MUX21(D,SI,SE) Q=DFF(DI,CLK) QN=INV1(Q) ;
+
+LATCHX{1,2}$ input(D,CLK) output(Q,QN) Q=LATCH(D,CLK) QN=INV1(Q) ;
+""".replace('$','{,_LVT,_HVT}'))
+"""The SAED 90nm educational technology library.
+It defines all cells except: negative-edge flip-flops, tri-state, latches, clock gating, level shifters
+"""
+
+
+SKY130 = TechLib(r"""
+$decap_{3,4,6,8,12}             ;
+$fill_{1,2,4,8}                 ;
+$tap_{1,2}                      ;
+$tapvgnd_1                      ;
+$tapvgnd2_1                     ;
+$tapvpwrvgnd_1                  ;
+$lpflow_decapkapwr_{3,4,6,8,12} ;
+$lpflow_bleeder_1  input(SHORT) ;
+
+$conb_1  output(HI,LO) HI=__const1__() LO=__const0__() ;
+$macro_sparecell  output(LO) LO=__const0__() ;
+
+$diode_2      input(DIODE) ;
+$probe_p_8    input(A) output(X) X=BUF1(A) ;
+$probec_p_8   input(A) output(X) X=BUF1(A) ;
+
+$inv_{1,2,4,6,8,12,16}                   input(A) output(Y) Y=INV1(A) ;
+$buf_{1,2,4,6,8,12,16}                   input(A) output(X) X=BUF1(A) ;
+$bufbuf_{8,16}                           input(A) output(X) X=BUF1(A) ;
+$bufinv_{8,16}                           input(A) output(Y) Y=INV1(A) ;
+$clkbuf_{1,2,4,8,16}                     input(A) output(X) X=BUF1(A) ;
+$clkinv_{1,2,4,8,16}                     input(A) output(Y) Y=INV1(A) ;
+$clkinvlp_{2,4}                          input(A) output(Y) Y=INV1(A) ;
+$clkdlybuf4s15_{1,2}                     input(A) output(X) X=BUF1(A) ;
+$clkdlybuf4s18_{1,2}                     input(A) output(X) X=BUF1(A) ;
+$clkdlybuf4s25_{1,2}                     input(A) output(X) X=BUF1(A) ;
+$clkdlybuf4s50_{1,2}                     input(A) output(X) X=BUF1(A) ;
+$dlygate4sd1_1                           input(A) output(X) X=BUF1(A) ;
+$dlygate4sd2_1                           input(A) output(X) X=BUF1(A) ;
+$dlygate4sd3_1                           input(A) output(X) X=BUF1(A) ;
+$dlymetal6s2s_1                          input(A) output(X) X=BUF1(A) ;
+$dlymetal6s4s_1                          input(A) output(X) X=BUF1(A) ;
+$dlymetal6s6s_1                          input(A) output(X) X=BUF1(A) ;
+$lpflow_clkbufkapwr_{1,2,4,8,16}         input(A) output(X) X=BUF1(A) ;
+$lpflow_clkinvkapwr_{1,2,4,8,16}         input(A) output(Y) Y=INV1(A) ;
+$lpflow_lsbuf_lh_hl_isowell_tap_{1,2,4}  input(A) output(X) X=BUF1(A) ;
+$lpflow_lsbuf_lh_isowell_4               input(A) output(X) X=BUF1(A) ;
+$lpflow_lsbuf_lh_isowell_tap_{1,2,4}     input(A) output(X) X=BUF1(A) ;
+
+$and2_{0,1,2,4}    input(A,B)          output(X) X=AND2(A,B)                                    ;
+$and2b_{1,2,4}     input(A_N,B)        output(X) AN=INV1(A_N) X=AND2(AN,B)                      ;
+$and3_{1,2,4}      input(A,B,C)        output(X) X=AND3(A,B,C)                                  ;
+$and3b_{1,2,4}     input(A_N,B,C)      output(X) AN=INV1(A_N) X=AND3(AN,B,C)                    ;
+$and4_{1,2,4}      input(A,B,C,D)      output(X) X=AND4(A,B,C,D)                                ;
+$and4b_{1,2,4}     input(A_N,B,C,D)    output(X) AN=INV1(A_N) X=AND4(AN,B,C,D)                  ;
+$and4bb_{1,2,4}    input(A_N,B_N,C,D)  output(X) AN=INV1(A_N) BN=INV1(B_N) X=AND4(AN,BN,C,D)    ;
+
+$or2_{0,1,2,4}     input(A,B)          output(X) X=OR2(A,B)                                     ;
+$or2b_{1,2,4}      input(A,B_N)        output(X) BN=INV1(B_N) X=OR2(A,BN)                       ;
+$or3_{1,2,4}       input(A,B,C)        output(X) X=OR3(A,B,C)                                   ;
+$or3b_{1,2,4}      input(A,B,C_N)      output(X) CN=INV1(C_N) X=OR3(A,B,CN)                     ;
+$or4_{1,2,4}       input(A,B,C,D)      output(X) X=OR4(A,B,C,D)                                 ;
+$or4b_{1,2,4}      input(A,B,C,D_N)    output(X) DN=INV1(D_N) X=OR4(A,B,C,DN)                   ;
+$or4bb_{1,2,4}     input(A,B,C_N,D_N)  output(X) CN=INV1(C_N) DN=INV1(D_N) X=OR4(A,B,CN,DN)     ;
+
+$nand2_{1,2,4,8}   input(A,B)          output(Y) Y=NAND2(A,B)                                   ;
+$nand2b_{1,2,4}    input(A_N,B)        output(Y) AN=INV1(A_N) Y=NAND2(AN,B)                     ;
+$nand3_{1,2,4}     input(A,B,C)        output(Y) Y=NAND3(A,B,C)                                 ;
+$nand3b_{1,2,4}    input(A_N,B,C)      output(Y) AN=INV1(A_N) Y=NAND3(AN,B,C)                   ;
+$nand4_{1,2,4}     input(A,B,C,D)      output(Y) Y=NAND4(A,B,C,D)                               ;
+$nand4b_{1,2,4}    input(A_N,B,C,D)    output(Y) AN=INV1(A_N) Y=NAND4(AN,B,C,D)                 ;
+$nand4bb_{1,2,4}   input(A_N,B_N,C,D)  output(Y) AN=INV1(A_N) BN=INV1(B_N) Y=NAND4(AN,BN,C,D)   ;
+
+$nor2_{1,2,4,8}    input(A,B)          output(Y) Y=NOR2(A,B)                                    ;
+$nor2b_{1,2,4}     input(A,B_N)        output(Y) BN=INV1(B_N) Y=NOR2(A,BN)                      ;
+$nor3_{1,2,4}      input(A,B,C)        output(Y) Y=NOR3(A,B,C)                                  ;
+$nor3b_{1,2,4}     input(A,B,C_N)      output(Y) CN=INV1(C_N) Y=NOR3(A,B,CN)                    ;
+$nor4_{1,2,4}      input(A,B,C,D)      output(Y) Y=NOR4(A,B,C,D)                                ;
+$nor4b_{1,2,4}     input(A,B,C,D_N)    output(Y) DN=INV1(D_N) Y=NOR4(A,B,C,DN)                  ;
+$nor4bb_{1,2,4}    input(A,B,C_N,D_N)  output(Y) CN=INV1(C_N) DN=INV1(D_N) Y=NOR4(A,B,CN,DN)    ;
+
+$xor2_{1,2,4}      input(A,B)   output(X) X=XOR2(A,B)   ;
+$xor3_{1,2,4}      input(A,B,C) output(X) X=XOR3(A,B,C) ;
+$xnor2_{1,2,4}     input(A,B)   output(Y) Y=XNOR2(A,B)  ;
+$xnor3_{1,2,4}     input(A,B,C) output(X) X=XNOR3(A,B,C) ;
+
+$maj3_{1,2,4}  input(A,B,C) output(X) AB=AND2(A,B) BC=AND2(B,C) AC=AND2(A,C) X=OR3(AB,BC,AC) ;
+
+$mux2_{1,2,4,8}  input(A0,A1,S)           output(X) X=MUX21(A0,A1,S)                                    ;
+$mux2i_{1,2,4}   input(A0,A1,S)           output(Y) M=MUX21(A0,A1,S) Y=INV1(M)                          ;
+$mux4_{1,2,4}    input(A0,A1,A2,A3,S0,S1) output(X) M0=MUX21(A0,A1,S0) M1=MUX21(A2,A3,S0) X=MUX21(M0,M1,S1) ;
+
+$ha_{1,2,4}      input(A,B)     output(COUT,SUM) SUM=XOR2(A,B) COUT=AND2(A,B)                           ;
+$fa_{1,2,4}      input(A,B,CIN) output(COUT,SUM) AB=XOR2(A,B) SUM=XOR2(AB,CIN) COUT=AO22(A,B,AB,CIN)    ;
+$fah_1           input(A,B,CI)  output(COUT,SUM) AB=XOR2(A,B) SUM=XOR2(AB,CI)  COUT=AO22(A,B,AB,CI)     ;
+$fahcin_1        input(A,B,CIN) output(COUT,SUM) AB=XOR2(A,B) SUM=XOR2(AB,CIN) COUT=AO22(A,B,AB,CIN)    ;
+$fahcon_1        input(A,B,CI)  output(COUT_N,SUM) AB=XOR2(A,B) SUM=XOR2(AB,CI) T=AO22(A,B,AB,CI) COUT_N=INV1(T) ;
+
+$a21o_{1,2,4}     input(A1,A2,B1)              output(X) X=AO21(A1,A2,B1)                               ;
+$a21bo_{1,2,4}    input(A1,A2,B1_N)            output(X) BN=INV1(B1_N) X=AO21(A1,A2,BN)                 ;
+$a22o_{1,2,4}     input(A1,A2,B1,B2)           output(X) X=AO22(A1,A2,B1,B2)                            ;
+$a2bb2o_{1,2,4}   input(A1_N,A2_N,B1,B2)       output(X) AN1=INV1(A1_N) AN2=INV1(A2_N) X=AO22(AN1,AN2,B1,B2) ;
+$a211o_{1,2,4}    input(A1,A2,B1,C1)           output(X) X=AO211(A1,A2,B1,C1)                           ;
+$a221o_{1,2,4}    input(A1,A2,B1,B2,C1)        output(X) T=AO22(A1,A2,B1,B2) X=OR2(T,C1)                ;
+$a31o_{1,2,4}     input(A1,A2,A3,B1)           output(X) AA=AND3(A1,A2,A3) X=OR2(AA,B1)                 ;
+$a311o_{1,2,4}    input(A1,A2,A3,B1,C1)        output(X) T=AND2(A1,A2) X=AO211(T,A3,B1,C1)              ;
+$a32o_{1,2,4}     input(A1,A2,A3,B1,B2)        output(X) T=AND2(A1,A2) X=AO22(T,A3,B1,B2)               ;
+$a41o_{1,2,4}     input(A1,A2,A3,A4,B1)        output(X) T1=AND2(A1,A2) T2=AND2(T1,A3) X=AO21(T2,A4,B1) ;
+$a2111o_{1,2,4}   input(A1,A2,B1,C1,D1)        output(X) T=AO211(A1,A2,B1,C1) X=OR2(T,D1)               ;
+
+$a21oi_{1,2,4}    input(A1,A2,B1)              output(Y) Y=AOI21(A1,A2,B1)                               ;
+$a21boi_{0,1,2,4} input(A1,A2,B1_N)            output(Y) BN=INV1(B1_N) Y=AOI21(A1,A2,BN)                 ;
+$a22oi_{1,2,4}    input(A1,A2,B1,B2)           output(Y) Y=AOI22(A1,A2,B1,B2)                            ;
+$a2bb2oi_{1,2,4}  input(A1_N,A2_N,B1,B2)       output(Y) AN1=INV1(A1_N) AN2=INV1(A2_N) Y=AOI22(AN1,AN2,B1,B2) ;
+$a211oi_{1,2,4}   input(A1,A2,B1,C1)           output(Y) Y=AOI211(A1,A2,B1,C1)                           ;
+$a221oi_{1,2,4}   input(A1,A2,B1,B2,C1)        output(Y) T=AO22(A1,A2,B1,B2) Y=NOR2(T,C1)                ;
+$a222oi_1         input(A1,A2,B1,B2,C1,C2)     output(Y) AB=AO22(A1,A2,B1,B2) Y=AOI21(C1,C2,AB)          ;
+$a31oi_{1,2,4}    input(A1,A2,A3,B1)           output(Y) AA=AND3(A1,A2,A3) Y=NOR2(AA,B1)                 ;
+$a311oi_{1,2,4}   input(A1,A2,A3,B1,C1)        output(Y) T=AND2(A1,A2) Y=AOI211(T,A3,B1,C1)              ;
+$a32oi_{1,2,4}    input(A1,A2,A3,B1,B2)        output(Y) T=AND2(A1,A2) Y=AOI22(T,A3,B1,B2)               ;
+$a41oi_{1,2,4}    input(A1,A2,A3,A4,B1)        output(Y) T1=AND2(A1,A2) T2=AND2(T1,A3) Y=AOI21(T2,A4,B1) ;
+$a2111oi_{0,1,2,4} input(A1,A2,B1,C1,D1)       output(Y) T=AO211(A1,A2,B1,C1) Y=NOR2(T,D1)               ;
+
+$o21a_{1,2,4}     input(A1,A2,B1)              output(X) X=OA21(A1,A2,B1)                              ;
+$o21ba_{1,2,4}    input(A1,A2,B1_N)            output(X) BN=INV1(B1_N) X=OA21(A1,A2,BN)                ;
+$o22a_{1,2,4}     input(A1,A2,B1,B2)           output(X) X=OA22(A1,A2,B1,B2)                           ;
+$o2bb2a_{1,2,4}   input(A1_N,A2_N,B1,B2)       output(X) T=NAND2(A1_N,A2_N) X=OA21(B1,B2,T)            ;
+$o211a_{1,2,4}    input(A1,A2,B1,C1)           output(X) X=OA211(A1,A2,B1,C1)                          ;
+$o221a_{1,2,4}    input(A1,A2,B1,B2,C1)        output(X) T=OA22(A1,A2,B1,B2) X=AND2(T,C1)              ;
+$o31a_{1,2,4}     input(A1,A2,A3,B1)           output(X) T=OR3(A1,A2,A3) X=AND2(T,B1)                  ;
+$o311a_{1,2,4}    input(A1,A2,A3,B1,C1)        output(X) T=OR2(A1,A2) X=OA211(T,A3,B1,C1)              ;
+$o32a_{1,2,4}     input(A1,A2,A3,B1,B2)        output(X) T=OR3(A1,A2,A3) X=OA21(B1,B2,T)               ;
+$o41a_{1,2,4}     input(A1,A2,A3,A4,B1)        output(X) T=OR4(A1,A2,A3,A4) X=AND2(T,B1)               ;
+$o2111a_{1,2,4}   input(A1,A2,B1,C1,D1)        output(X) T=OA211(A1,A2,B1,C1) X=AND2(T,D1)             ;
+
+$o21ai_{0,1,2,4}  input(A1,A2,B1)              output(Y) Y=OAI21(A1,A2,B1)                             ;
+$o21bai_{1,2,4}   input(A1,A2,B1_N)            output(Y) BN=INV1(B1_N) Y=OAI21(A1,A2,BN)               ;
+$o22ai_{1,2,4}    input(A1,A2,B1,B2)           output(Y) Y=OAI22(A1,A2,B1,B2)                          ;
+$o2bb2ai_{1,2,4}  input(A1_N,A2_N,B1,B2)       output(Y) T=NAND2(A1_N,A2_N) Y=OAI21(B1,B2,T)           ;
+$o211ai_{1,2,4}   input(A1,A2,B1,C1)           output(Y) Y=OAI211(A1,A2,B1,C1)                         ;
+$o221ai_{1,2,4}   input(A1,A2,B1,B2,C1)        output(Y) T=OA22(A1,A2,B1,B2) Y=NAND2(T,C1)             ;
+$o31ai_{1,2,4}    input(A1,A2,A3,B1)           output(Y) T=OR3(A1,A2,A3) Y=NAND2(T,B1)                 ;
+$o311ai_{0,1,2,4} input(A1,A2,A3,B1,C1)        output(Y) T=OR2(A1,A2) Y=OAI211(T,A3,B1,C1)             ;
+$o32ai_{1,2,4}    input(A1,A2,A3,B1,B2)        output(Y) T=OR3(A1,A2,A3) Y=OAI21(B1,B2,T)              ;
+$o41ai_{1,2,4}    input(A1,A2,A3,A4,B1)        output(Y) T=OR4(A1,A2,A3,A4) Y=NAND2(T,B1)              ;
+$o2111ai_{1,2,4}  input(A1,A2,B1,C1,D1)        output(Y) T=OA211(A1,A2,B1,C1) Y=NAND2(T,D1)            ;
+
+$dfxtp_{1,2,4}    input(CLK,D)                 output(Q)     Q=DFF(D,CLK)                                                              ;
+$dfxbp_{1,2}      input(CLK,D)                 output(Q,Q_N) Q=DFF(D,CLK) Q_N=INV1(Q)                                                  ;
+$dfrtp_{1,2,4}    input(CLK,D,RESET_B)         output(Q)     DR=AND2(D,RESET_B) Q=DFF(DR,CLK)                                          ;
+$dfrbp_{1,2}      input(CLK,D,RESET_B)         output(Q,Q_N) DR=AND2(D,RESET_B) Q=DFF(DR,CLK) Q_N=INV1(Q)                              ;
+$dfstp_{1,2,4}    input(CLK,D,SET_B)           output(Q)     S=INV1(SET_B) DS=OR2(D,S) Q=DFF(DS,CLK)                                   ;
+$dfsbp_{1,2}      input(CLK,D,SET_B)           output(Q,Q_N) S=INV1(SET_B) DS=OR2(D,S) Q=DFF(DS,CLK) Q_N=INV1(Q)                       ;
+$dfbbp_1          input(CLK,D,SET_B,RESET_B)   output(Q,Q_N) DR=AND2(D,RESET_B) S=INV1(SET_B) DRS=OR2(DR,S) Q=DFF(DRS,CLK) Q_N=INV1(Q) ;
+$dfrtn_1          input(CLK_N,D,RESET_B)       output(Q)     CLKN=INV1(CLK_N) DR=AND2(D,RESET_B) Q=DFF(DR,CLKN)                        ;
+$dfbbn_{1,2}      input(CLK_N,D,SET_B,RESET_B) output(Q,Q_N) CLKN=INV1(CLK_N) DR=AND2(D,RESET_B) S=INV1(SET_B) DRS=OR2(DR,S) Q=DFF(DRS,CLKN) Q_N=INV1(Q) ;
+
+$sdfxtp_{1,2,4}   input(CLK,D,SCD,SCE)                 output(Q)     DI=MUX21(D,SCD,SCE) Q=DFF(DI,CLK)                                        ;
+$sdfxbp_{1,2}     input(CLK,D,SCD,SCE)                 output(Q,Q_N) DI=MUX21(D,SCD,SCE) Q=DFF(DI,CLK) Q_N=INV1(Q)                            ;
+$sdfrtp_{1,2,4}   input(CLK,D,SCD,SCE,RESET_B)         output(Q)     DR=AND2(D,RESET_B) DI=MUX21(DR,SCD,SCE) Q=DFF(DI,CLK)                    ;
+$sdfrbp_{1,2}     input(CLK,D,SCD,SCE,RESET_B)         output(Q,Q_N) DR=AND2(D,RESET_B) DI=MUX21(DR,SCD,SCE) Q=DFF(DI,CLK) Q_N=INV1(Q)        ;
+$sdfstp_{1,2,4}   input(CLK,D,SCD,SCE,SET_B)           output(Q)     S=INV1(SET_B) DS=OR2(D,S) DI=MUX21(DS,SCD,SCE) Q=DFF(DI,CLK)             ;
+$sdfsbp_{1,2}     input(CLK,D,SCD,SCE,SET_B)           output(Q,Q_N) S=INV1(SET_B) DS=OR2(D,S) DI=MUX21(DS,SCD,SCE) Q=DFF(DI,CLK) Q_N=INV1(Q) ;
+$sdfbbp_1         input(CLK,D,SCD,SCE,SET_B,RESET_B)   output(Q,Q_N) DR=AND2(D,RESET_B) S=INV1(SET_B) DRS=OR2(DR,S) DI=MUX21(DRS,SCD,SCE) Q=DFF(DI,CLK) Q_N=INV1(Q) ;
+$sdfrtn_1         input(CLK_N,D,SCD,SCE,RESET_B)       output(Q)     CLKN=INV1(CLK_N) DR=AND2(D,RESET_B) DI=MUX21(DR,SCD,SCE) Q=DFF(DI,CLKN) ;
+$sdfbbn_{1,2}     input(CLK_N,D,SCD,SCE,SET_B,RESET_B) output(Q,Q_N) CLKN=INV1(CLK_N) DR=AND2(D,RESET_B) S=INV1(SET_B) DRS=OR2(DR,S) DI=MUX21(DRS,SCD,SCE) Q=DFF(DI,CLKN) Q_N=INV1(Q) ;
+
+$edfxtp_1         input(CLK,D,DE)                    output(Q)     GCK=AND2(CLK,DE) Q=DFF(D,GCK)                                           ;
+$edfxbp_1         input(CLK,D,DE)                    output(Q,Q_N) GCK=AND2(CLK,DE) Q=DFF(D,GCK) Q_N=INV1(Q)                               ;
+$sedfxtp_{1,2,4}  input(CLK,D,DE,SCD,SCE)            output(Q)     DI=MUX21(D,SCD,SCE) GCK=AND2(CLK,DE) Q=DFF(DI,GCK)                      ;
+$sedfxbp_{1,2}    input(CLK,D,DE,SCD,SCE)            output(Q,Q_N) DI=MUX21(D,SCD,SCE) GCK=AND2(CLK,DE) Q=DFF(DI,GCK) Q_N=INV1(Q)          ;
+
+$dlxtp_1          input(D,GATE)                      output(Q)     Q=LATCH(D,GATE)                                                         ;
+$dlxtn_{1,2,4}    input(D,GATE_N)                    output(Q)     GN=INV1(GATE_N) Q=LATCH(D,GN)                                           ;
+$dlxbp_1          input(D,GATE)                      output(Q,Q_N) Q=LATCH(D,GATE) Q_N=INV1(Q)                                             ;
+$dlxbn_{1,2}      input(D,GATE_N)                    output(Q,Q_N) GN=INV1(GATE_N) Q=LATCH(D,GN) Q_N=INV1(Q)                               ;
+$dlrtp_{1,2,4}    input(D,GATE,RESET_B)              output(Q)     DR=AND2(D,RESET_B) Q=LATCH(DR,GATE)                                     ;
+$dlrtn_{1,2,4}    input(D,GATE_N,RESET_B)            output(Q)     GN=INV1(GATE_N) DR=AND2(D,RESET_B) Q=LATCH(DR,GN)                       ;
+$dlrbp_{1,2}      input(D,GATE,RESET_B)              output(Q,Q_N) DR=AND2(D,RESET_B) Q=LATCH(DR,GATE) Q_N=INV1(Q)                         ;
+$dlrbn_{1,2}      input(D,GATE_N,RESET_B)            output(Q,Q_N) GN=INV1(GATE_N) DR=AND2(D,RESET_B) Q=LATCH(DR,GN) Q_N=INV1(Q)           ;
+
+$dlclkp_{1,2,4}   input(CLK,GATE)                    output(GCLK) GCLK=AND2(CLK,GATE)                                                      ;
+$sdlclkp_{1,2,4}  input(CLK,GATE,SCE)                output(GCLK) G=OR2(GATE,SCE) GCLK=AND2(CLK,G)                                         ;
+
+$ebufn_{1,2,4,8}    input(A,TE_B) output(Z) TE=INV1(TE_B) Z=AND2(A,TE)               ;
+$einvn_{0,1,2,4,8}  input(A,TE_B) output(Z) TE=INV1(TE_B) ZI=INV1(A) Z=AND2(ZI,TE)   ;
+$einvp_{1,2,4,8}    input(A,TE)   output(Z) ZI=INV1(A) Z=AND2(ZI,TE)                 ;
+
+$lpflow_inputiso0n_1  input(A,SLEEP_B) output(X) X=AND2(A,SLEEP_B)                   ;
+$lpflow_inputiso0p_1  input(A,SLEEP)   output(X) SB=INV1(SLEEP) X=AND2(A,SB)         ;
+$lpflow_inputiso1n_1  input(A,SLEEP_B) output(X) X=AND2(A,SLEEP_B)                   ;
+$lpflow_inputiso1p_1  input(A,SLEEP)   output(X) SB=INV1(SLEEP) X=AND2(A,SB)         ;
+$lpflow_isobufsrc_{1,2,4,8,16}   input(A,SLEEP) output(X) AI=INV1(A) X=OR2(AI,SLEEP) ;
+$lpflow_isobufsrckapwr_16        input(A,SLEEP) output(X) AI=INV1(A) X=OR2(AI,SLEEP) ;
+$lpflow_inputisolatch_1  input(D,SLEEP_B) output(Q) Q=LATCH(D,SLEEP_B)               ;
+""".replace('$','sky130_fd_sc_hd__'))
+"""SkyWater 130nm High Density Digital Standard Cells (skywater-pdk-libs-sky130_fd_sc_hd).
+"""
+
+techlib_by_name = {k: v for k, v in globals().items() if isinstance(v, TechLib)}

src/kyupy/vcd.py

@@ -0,0 +1,218 @@
+"""A simple parser for Verilog Change Dump (VCD) files.
+
+This parser loads the changes into a ndarray of logic values.
+Axis 0 is the time step, axis 1 is the variable data.
+All variable values are flattened. Offsets are stored in each variable metadata.
+"""
+
+from collections import namedtuple
+from dataclasses import dataclass, field
+
+from lark import Lark, Transformer
+import numpy as np
+
+from . import readtext, logic
+
+
+@dataclass
+class Var:
+    type: str
+    width: int
+    idcode: str
+    reference: str
+    scope: 'Scope'
+    locs: list[int] = field(default_factory=list)
+
+
+class Scope:
+    def __init__(self, parent, name='', type='module') -> None:
+        self.name = name
+        self.type = type
+        self.parent_scope = parent
+        self.sub_scopes : list[Scope] = []
+        self.vars : list = []
+
+    @property
+    def path(self) -> str:
+        parts = []
+        s = self
+        while s is not None and s.name:
+            parts.append(s.name)
+            s = s.parent_scope
+        return '/'.join(reversed(parts))
+
+    def __repr__(self) -> str:
+        return self.path
+
+
+class VcdHeader:
+    def __init__(self) -> None:
+        self.comment: str = ''
+        self.date: str = ''
+        self.timescale: str = ''
+        self.version: str = ''
+        self.root_scope = Scope(None)
+        self.current_scope = self.root_scope
+
+    def scope(self, name, type):
+        new_scope = Scope(self.current_scope, name, type)
+        self.current_scope.sub_scopes.append(new_scope)
+        self.current_scope = new_scope
+
+    def upscope(self):
+        assert self.current_scope.parent_scope is not None, "root scope has no parent"
+        self.current_scope = self.current_scope.parent_scope
+
+    def add_var(self, type: str, width: int, idcode: str, reference: str):
+        self.current_scope.vars.append(Var(type, width, idcode, reference, self.current_scope))
+
+    def __repr__(self):
+        lines = [
+            'VcdHeader(',
+            f'  comment   = {self.comment!r}',
+            f'  date      = {self.date!r}',
+            f'  timescale = {self.timescale!r}',
+            f'  version   = {self.version!r}',
+            '  scopes:',
+        ]
+        def fmt_scope(scope, indent=4):
+            lines.append(f'{"  " * (indent // 2)}{scope.type}:{scope.name} ({len(scope.vars)} vars)')
+            for sub in scope.sub_scopes:
+                fmt_scope(sub, indent + 2)
+        for sub in self.root_scope.sub_scopes:
+            fmt_scope(sub)
+        lines.append(')')
+        return '\n'.join(lines)
+
+
+class VcdVarMap:
+    def __init__(self, header: VcdHeader, var_locs = lambda _: []) -> None:
+        self.var_list: list[Var] = []
+        def collect(scope):
+            for v in scope.vars:
+                v.locs = var_locs(v)
+                if v.locs is not None and len(v.locs) > 0:
+                    assert len(v.locs) == v.width, f'{v.reference}: len(locs)={len(v.locs)} != width={v.width}'
+                    self.var_list.append(v)
+            for sub in scope.sub_scopes:
+                collect(sub)
+        collect(header.root_scope)
+        self.total_width = max((max(v.locs) for v in self.var_list), default=-1) + 1
+        self.idcode2var = {var.idcode: var for var in self.var_list}
+
+
+class VcdData:
+    def __init__(self, var_map: VcdVarMap, steps, data):
+        self.var_map = var_map
+        self.steps = steps
+        self.data = data
+    def __repr__(self) -> str:
+        return f'{{variables: {len(self.var_map.var_list)}, bits: {self.data.shape[0]}, steps: {self.data.shape[1]}}}'
+
+
+class VcdHeaderTransformer(Transformer):
+    def __init__(self):
+        super().__init__()
+        self.header = VcdHeader()
+
+    def comment(self, args): self.header.comment = args[0].value.strip()
+    def date(self, args): self.header.date = args[0].value.strip()
+    def timescale(self, args): self.header.timescale = args[0].value.strip()
+    def version(self, args): self.header.version = args[0].value.strip()
+
+    def scope(self, args):
+        type, name = args
+        self.header.scope(name.value, type.value)
+
+    def upscope(self, args):
+        self.header.upscope()
+
+    def var(self, args):
+        type, width, idcode, reference = args
+        reference = reference.value.strip()
+        width = int(width)
+        self.header.add_var(type.value, width, idcode.value, reference)
+
+    def start(self, args):
+        return self.header
+        #return tuple(args) if len(args) > 1 else (args[0], None)
+
+
+GRAMMAR = r"""
+    start: (_declaration_command)*
+    _declaration_command: ( comment | date | scope | timescale | upscope | var | version ) "$end"
+    comment: "$comment" TEXT?
+    date: "$date" TEXT
+    scope: "$scope" /[^\s]+/ /[^\s]+/
+    timescale: "$timescale" TEXT
+    upscope: "$upscope"
+    var: "$var" /[^\s]+/ /[^\s]+/ /[^\s]+/ TEXT
+    version: "$version" TEXT
+    TEXT: /(?:(?!\$end)[\s\S])+/
+    %ignore ( /\r?\n/ )+
+    %ignore /[\t \f]+/
+    """
+
+
+def load(file, var_locs = lambda _: [], step_filter = lambda *_: True, s_len = None) -> VcdData:
+    """Parses the contents of ``file`` as Verilog Change Dump (VCD).
+
+    :param file: A file name or a file handle. Files with `.gz`-suffix are decompressed on-the-fly.
+    :param var_locs: A callback ``(var) -> list[int]`` mapping each variable to ndarray column indices. Empty list drops the variable.
+    :param step_filter: A callback ``(time, values, var_map) -> bool`` to select which timesteps to include.
+    :param s_len: length of the first axis of returned pattern data.
+    :return: A VcdData object with metadata and an ndarray with all values.
+    """
+    vcd = readtext(file)
+    header_size = vcd.find('$enddefinitions')
+    assert header_size > 0, "invalid VCD file: end of header not found"
+    vcd_header_str = vcd[:header_size]
+    vcd_header : VcdHeader = Lark(GRAMMAR, parser="lalr", lexer='contextual', transformer=VcdHeaderTransformer()).parse(vcd_header_str) # type: ignore
+    vcd_data = vcd[header_size:].splitlines()
+    var_map = VcdVarMap(vcd_header, var_locs)
+    width = s_len if s_len is not None else var_map.total_width
+
+    chunk_size = 10240
+    chunks = []
+    chunk = np.full((chunk_size, width), logic.UNASSIGNED, dtype=np.uint8)
+
+    _val_map = {'0': logic.ZERO, '1': logic.ONE,
+                'x': logic.UNKNOWN, 'X': logic.UNKNOWN,
+                'z': logic.UNASSIGNED, 'Z': logic.UNASSIGNED}
+    _pad_char = {'0': '0', '1': '0', 'x': 'x', 'X': 'x', 'z': 'z', 'Z': 'z'}
+
+    current_time = None
+    steps = []
+    step_idx = 0
+    for line in vcd_data:
+        if not line or line[0] == '$':
+            continue
+        if line[0] == '#':
+            if step_filter(current_time, chunk[step_idx], var_map):
+                step_idx += 1
+                steps.append(current_time)
+                if step_idx >= chunk_size:
+                    chunks.append(chunk)
+                    chunk = np.empty((chunk_size, width), dtype=np.uint8)
+                    chunk[0] = chunks[-1][-1]
+                    step_idx = 0
+                else:
+                    chunk[step_idx] = chunk[step_idx - 1]
+            current_time = int(line[1:])
+            continue
+        if line[0] in ('b', 'B'):
+            value_str, idcode = line[1:].split(None, 1)
+        else:
+            value_str, idcode = line[0], line[1:]
+        var = var_map.idcode2var.get(idcode)
+        if var is None:
+            continue
+        if len(value_str) < var.width:
+            value_str = _pad_char.get(value_str[0], '0') * (var.width - len(value_str)) + value_str
+        for i, c in enumerate(value_str):
+            if var.locs[i] >= 0:
+                chunk[step_idx, var.locs[i]] = _val_map.get(c, logic.UNKNOWN)
+
+    chunks.append(chunk[:step_idx])
+    data = np.concatenate(chunks, axis=0).T
+    return VcdData(var_map, steps, data)

src/kyupy/verilog.py

@@ -1,173 +1,294 @@
+"""A simple and incomplete parser for Verilog files.
+
+The main purpose of this parser is to load synthesized, non-hierarchical (flat) gate-level netlists.
+It supports only a subset of Verilog.
+"""
+
 from collections import namedtuple
-import gzip
 
-from lark import Lark, Transformer
+from lark import Lark, Transformer, Tree
 
+from . import log, readtext
 from .circuit import Circuit, Node, Line
-from .saed import pin_index, pin_is_output
+from .techlib import KYUPY
 
 Instantiation = namedtuple('Instantiation', ['type', 'name', 'pins'])
 
 
 class SignalDeclaration:
-    
-    def __init__(self, kind, tokens):
+
+    def __init__(self, kind, name, rnge=None):
         self.left = None
         self.right = None
         self.kind = kind
-        if len(tokens.children) == 1:
-            self.basename = tokens.children[0]
-        else:
-            self.basename = tokens.children[2]
-            self.left = int(tokens.children[0].value)
-            self.right = int(tokens.children[1].value)
-    
+        self.basename = name
+        self.rnge = rnge
+
     @property
     def names(self):
-        if self.left is None:
+        if self.rnge is None:
             return [self.basename]
-        if self.left <= self.right:
-            return [f'{self.basename}[{i}]' for i in range(self.left, self.right + 1)]
-        else:
-            return [f'{self.basename}[{i}]' for i in range(self.left, self.right - 1, -1)]
-        
+        return [f'{self.basename}[{i}]' for i in self.rnge]
+
     def __repr__(self):
-        return f"{self.kind}:{self.basename}[{self.left}:{self.right}]"
+        return f"{self.kind}:{self.basename}[{self.rnge}]"
 
 
 class VerilogTransformer(Transformer):
-    def __init__(self, branchforks=False):
+    def __init__(self, branchforks=False, tlib=KYUPY):
         super().__init__()
-        self._signal_declarations = {}
         self.branchforks = branchforks
+        self.tlib = tlib
 
     @staticmethod
     def name(args):
         s = args[0].value
-        if s[0] == '\\':
-            s = s[1:-1]
-        return s
+        return s[1:-1] if s[0] == '\\' else s
+
+    @staticmethod
+    def namedpin(args):
+        return tuple(args) if len(args) > 1 else (args[0], None)
 
     @staticmethod
     def instantiation(args):
-        return Instantiation(args[0], args[1],
-                             dict([(pin.children[0], pin.children[1]) for pin in args[2:]]))
-       
-    def input(self, args):
-        for sd in [SignalDeclaration('input', signal) for signal in args]:
-            self._signal_declarations[sd.basename] = sd
-    
-    def inout(self, args):
-        for sd in [SignalDeclaration('input', signal) for signal in args]:  # just treat as input
-            self._signal_declarations[sd.basename] = sd
-    
-    def output(self, args):
-        for sd in [SignalDeclaration('output', signal) for signal in args]:
-            self._signal_declarations[sd.basename] = sd
-            
-    def wire(self, args):
-        for sd in [SignalDeclaration('wire', signal) for signal in args]:
-            self._signal_declarations[sd.basename] = sd
-                
+        pinmap = {}
+        if args[2] is not None:
+            for idx, pin in enumerate(args[2:]):
+                p = pin.children[0]
+                if isinstance(p, tuple):  # named pin
+                    if p[1] is not None:
+                        pinmap[p[0]] = p[1]
+                else:  # unnamed pin
+                    pinmap[idx] = p
+        return Instantiation(args[0], args[1], pinmap)
+
+    @staticmethod
+    def simple_dff(args):
+        return Instantiation('DFF', args[1], {'d': args[2], 'clk': args[0], 'q': args[1]} )
+
+    def range(self, args):
+        left = int(args[0].value)
+        right = int(args[1].value) if len(args) > 1 else left
+        return range(left, right+1) if left <= right else range(left, right-1, -1)
+
+    def sigsel(self, args):
+        if len(args) > 1 and isinstance(args[1], range):
+            l = [f'{args[0]}[{i}]' for i in args[1]]
+            return l if len(l) > 1 else l[0]
+        elif "'" in args[0]:
+            width, rest = args[0].split("'")
+            width = int(width)
+            base, const = rest[0], rest[1:].replace('x','0')
+            const = int(const, {'b': 2, 'd':10, 'h':16}[base.lower()])
+            l = []
+            for _ in range(width):
+                l.insert(0, "1'b1" if (const & 1) else "1'b0")
+                const >>= 1
+            return l if len(l) > 1 else l[0]
+        else:
+            return args[0]
+
+    def concat(self, args):
+        sigs = []
+        for a in args:
+            if isinstance(a, list):
+                sigs += a
+            else:
+                sigs.append(a)
+        return sigs
+
+    def ternaryif(self, args):
+        sel = args[0]
+        ctrue = args[1]
+        cfalse = args[2]
+        log.warn(f"FIXME: not implemented: ternary if {args[0]} {args[1]}")
+        return args[1]
+
+    def declaration(self, kind, args):
+        rnge = None
+        if isinstance(args[0], range):
+            rnge = args[0]
+            args = args[1:]
+        return [SignalDeclaration(kind, signal, rnge) for signal in args]
+
+    def input(self, args): return self.declaration("input", args)
+    def output(self, args): return self.declaration("output", args)
+    def inout(self, args): return self.declaration("input", args)  # just treat as input
+    def wire(self, args): return self.declaration("wire", args)
+    def reg(self, args): pass #return self.declaration("wire", args)  # treat as wire
+
     def module(self, args):
         c = Circuit(args[0])
         positions = {}
         pos = 0
+        const_count = 0
+        sig_decls = {}
+        for decls in args[2:]:  # pass 0: collect signal declarations
+            if isinstance(decls, list):
+                if len(decls) > 0 and isinstance(decls[0], SignalDeclaration):
+                    for decl in decls:
+                        if decl.basename not in sig_decls or sig_decls[decl.basename].kind == 'wire':
+                            sig_decls[decl.basename] = decl
         for intf_sig in args[1].children:
-            for name in self._signal_declarations[intf_sig].names:
+            for name in sig_decls[intf_sig].names:
                 positions[name] = pos
                 pos += 1
         assignments = []
         for stmt in args[2:]:  # pass 1: instantiate cells and driven signals
-            if type(stmt) is Instantiation:
+            if isinstance(stmt, Instantiation):
+                if stmt.type not in self.tlib.cells:
+                    log.warn(f'Ignoring cell of unknown kind "{stmt.type}"')
+                    continue
                 n = Node(c, stmt.name, kind=stmt.type)
                 for p, s in stmt.pins.items():
-                    if pin_is_output(n.kind, p):
-                        Line(c, (n, pin_index(stmt.type, p)), Node(c, s))
-            elif stmt is not None and stmt.data == 'assign':
+                    if self.tlib.pin_is_output(n.kind, p):
+                        if s in sig_decls:
+                            s = sig_decls[s].names
+                            if isinstance(s, list) and len(s) == 1:
+                                s = s[0]
+                        Line(c, (n, self.tlib.pin_index(stmt.type, p)), Node(c, s))
+            elif hasattr(stmt, 'data') and stmt.data == 'assign':
                 assignments.append((stmt.children[0], stmt.children[1]))
-        for sd in self._signal_declarations.values():
+        for sd in sig_decls.values():
             if sd.kind == 'output' or sd.kind == 'input':
                 for name in sd.names:
                     n = Node(c, name, kind=sd.kind)
                     if name in positions:
-                        c.interface[positions[name]] = n
+                        c.io_nodes[positions[name]] = n
                     if sd.kind == 'input':
                         Line(c, n, Node(c, name))
-        for s1, s2 in assignments:  # pass 1.5: process signal assignments
-            if s1 in c.forks:
-                assert s2 not in c.forks, 'assignment between two driven signals'
-                Line(c, c.forks[s1], Node(c, s2))
-            elif s2 in c.forks:
-                assert s1 not in c.forks, 'assignment between two driven signals'
-                Line(c, c.forks[s2], Node(c, s1))
+        deferred_assignments = set()
+        ignored = 0
+        while len(assignments) > 0:
+            more_assignments = []
+            for target, source in assignments:  # pass 1.5: process signal assignments
+                target_sigs = []
+                if not isinstance(target, list): target = [target]
+                for s in target:
+                    if s in sig_decls:
+                        target_sigs += sig_decls[s].names
+                    else:
+                        target_sigs.append(s)
+                source_sigs = []
+                if not isinstance(source, list): source = [source]
+                for s in source:
+                    if s in sig_decls:
+                        source_sigs += sig_decls[s].names
+                    else:
+                        source_sigs.append(s)
+                for t, s in zip(target_sigs, source_sigs):
+                    if t in c.forks:
+                        assert s not in c.forks, f'assignment between two driven signals: source={s} target={t}'
+                        Line(c, c.forks[t], Node(c, s))
+                    elif s in c.forks:
+                        assert t not in c.forks, 'assignment between two driven signals'
+                        Line(c, c.forks[s], Node(c, t))
+                    elif s.startswith("1'b"):
+                        cnode = Node(c, f'__const{s[3]}_{const_count}__', f'__const{s[3]}__')
+                        const_count += 1
+                        Line(c, cnode, Node(c, t))
+                    else:
+                        if (t, s) in deferred_assignments:
+                            #log.info(f'ignoring: assign {t} = {s}')
+                            ignored += 1
+                        else:
+                            more_assignments.append((t, s))
+                            deferred_assignments.add((t, s)) 
+            assignments = more_assignments
+        if ignored > 0:
+            log.warn(f'ignored {ignored} assignments')
         for stmt in args[2:]:  # pass 2: connect signals to readers
-            if type(stmt) is Instantiation:
+            if isinstance(stmt, Instantiation):
                 for p, s in stmt.pins.items():
+                    if stmt.name not in c.cells: continue
                     n = c.cells[stmt.name]
-                    if pin_is_output(n.kind, p): continue
+                    if self.tlib.pin_is_output(n.kind, p): continue
                     if s.startswith("1'b"):
-                        const = f'__const{s[3]}__'
-                        if const not in c.cells:
-                            Line(c, Node(c, const, const), Node(c, s))
+                        cname = f'__const{s[3]}_{const_count}__'
+                        cnode = Node(c, cname, f'__const{s[3]}__')
+                        const_count += 1
+                        s = cname
+                        Line(c, cnode, Node(c, s))
+                    if s not in c.forks:
+                        if f'{s}[0]' in c.forks:  # actually a 1-bit bus?
+                            s = f'{s}[0]'
+                        else:
+                            log.warn(f'Signal not driven: {s}')
+                            Node(c, s)  # generate fork here
                     fork = c.forks[s]
                     if self.branchforks:
                         branchfork = Node(c, fork.name + "~" + n.name + "/" + p)
                         Line(c, fork, branchfork)
                         fork = branchfork
-                    Line(c, fork, (n, pin_index(stmt.type, p)))
-        for sd in self._signal_declarations.values():
+                    Line(c, fork, (n, self.tlib.pin_index(stmt.type, p)))
+        for sd in sig_decls.values():
             if sd.kind == 'output':
                 for name in sd.names:
+                    if name not in c.forks:
+                        if f'{name}[0]' in c.forks:  # actually a 1-bit bus?
+                            name = f'{name}[0]'
+                        else:
+                            log.warn(f'Output not driven: {name}')
+                            continue
                     Line(c, c.forks[name], c.cells[name])
         return c
 
     @staticmethod
-    def start(args):
-        if len(args) == 1:
-            return args[0]
-        else:
-            return args
+    def start(args): return args[0] if len(args) == 1 else args
 
 
-grammar = """
+GRAMMAR = r"""
     start: (module)*
     module: "module" name parameters ";" (_statement)* "endmodule"
-    parameters: "(" [ name ( "," name )* ] ")"
-    _statement: input | output | inout | tri | wire | assign | instantiation
-    input: "input" signal ( "," signal )* ";"
-    output: "output" signal ( "," signal )* ";"
-    inout: "inout" signal ( "," signal )* ";"
-    tri: "tri" name ";"
-    wire: "wire" signal ( "," signal )* ";"
-    assign: "assign" name "=" name ";"
+    parameters: "(" [ _namelist ] ")"
+    _statement: input | output | inout | tri | wire | reg | assign | instantiation | simple_dff
+    input: "input" range? _namelist ";"
+    output: "output" range? _namelist ";"
+    inout: "inout" range? _namelist ";"
+    tri: "tri" range? _namelist ";"
+    wire: "wire" range? _namelist ";"
+    reg: "reg" range? _namelist ";"
+    assign: "assign" sigsel "=" sigsel ";"
     instantiation: name name "(" [ pin ( "," pin )* ] ")" ";"
-    pin: "." name "(" name ")"
-    signal: ( name | "[" /[0-9]+/ ":" /[0-9]+/ "]" name )
-
-    name: ( /[a-z_][a-z0-9_\\[\\]]*/i | /\\\\[^\\t \\r\\n]+[\\t \\r\\n](\\[[0-9]+\\])?/i | /1'b0/i | /1'b1/i )
-    COMMENT: "//" /[^\\n]*/
-    %ignore ( /\\r?\\n/ | COMMENT )+
-    %ignore /[\\t \\f]+/
+    simple_dff: "always" "@" "(" "posedge" name ")" sigsel "<=" sigsel ";"
+    pin: namedpin | sigsel
+    namedpin: "." name "(" sigsel? ")"
+    range: "[" /[0-9]+/ (":" /[0-9]+/)? "]"
+    sigsel: name range? | concat | ternaryif
+    concat: "{" sigsel ( "," sigsel )*  "}"
+    ternaryif: sigsel "?" sigsel ":" sigsel
+    _namelist: name ( "," name )*
+    name: ( /[a-z_][a-z0-9_]*/i | /\\[^\t \r\n]+[\t \r\n]/i | /[0-9]+'[bdh][x0-9a-f]+/i )
+    %import common.NEWLINE
+    COMMENT: /\/\*(\*(?!\/)|[^*])*\*\// | /\(\*(\*(?!\))|[^*])*\*\)/ |  "//" /(.)*/ NEWLINE
+    %ignore ( /\r?\n/ | COMMENT )+
+    %ignore /[\t \f]+/
     """
 
 
-def loads(s, *, branchforks=False):
-    return Lark(grammar, parser="lalr", transformer=VerilogTransformer(branchforks)).parse(s)
+def parse(text, tlib=KYUPY, branchforks=False) -> Circuit:
+    """Parses the given ``text`` as Verilog code.
+
+    :param text: A string with Verilog code.
+    :param tlib: A technology library object that defines all known cells.
+    :type tlib: :py:class:`~kyupy.techlib.TechLib`
+    :param branchforks: If set to ``True``, the returned circuit will include additional `forks` on each fanout branch.
+        These forks are needed to correctly annotate interconnect delays
+        (see :py:func:`~kyupy.sdf.DelayFile.interconnects()`).
+    :return: A :py:class:`~kyupy.circuit.Circuit` object.
+    """
+    return Lark(GRAMMAR, parser="lalr", transformer=VerilogTransformer(branchforks, tlib)).parse(text) # type: ignore
 
 
-def load(fp, *, branchforks=False):
-    return loads(fp.read(), branchforks=branchforks)
+def load(file, tlib=KYUPY, branchforks=False):
+    """Parses the contents of ``file`` as Verilog code.
 
-
-def parse(verilog, branchforks=False):
-    if '\n' not in str(verilog):  # One line?: Assuming it is a file name.
-        if str(verilog).endswith('.gz'):
-            with gzip.open(verilog, 'rt') as f:
-                text = f.read()
-        else:
-            with open(verilog, 'r') as f:
-                text = f.read()
-    else:
-        text = str(verilog)
-    return loads(text, branchforks=branchforks)
+    :param file: A file name or a file handle. Files with `.gz`-suffix are decompressed on-the-fly.
+    :param tlib: A technology library object that defines all known cells.
+    :type tlib: :py:class:`~kyupy.techlib.TechLib`
+    :param branchforks: If set to ``True``, the returned circuit will include additional `forks` on each fanout branch.
+        These forks are needed to correctly annotate interconnect delays
+        (see :py:func:`~kyupy.sdf.DelayFile.interconnects()`).
+    :return: A :py:class:`~kyupy.circuit.Circuit` object.
+    """
+    return parse(readtext(file), tlib, branchforks)

src/kyupy/wave_sim_cuda.py

@@ -1,317 +0,0 @@
-import numpy as np
-import math
-from .wave_sim import WaveSim
-from . import cuda
-
-TMAX = np.float32(2 ** 127)  # almost np.PINF for 32-bit floating point values
-TMAX_OVL = np.float32(1.1 * 2 ** 127)  # almost np.PINF with overflow mark
-TMIN = np.float32(-2 ** 127)  # almost np.NINF for 32-bit floating point values
-
-
-class WaveSimCuda(WaveSim):
-    def __init__(self, circuit, timing, sims=8, wavecaps=16, strip_forks=False, keep_waveforms=True):
-        super().__init__(circuit, timing, sims, wavecaps, strip_forks, keep_waveforms)
-
-        self.tdata = np.zeros((len(self.interface), 3, (sims - 1) // 8 + 1), dtype='uint8')
-
-        self.d_state = cuda.to_device(self.state)
-        self.d_sat = cuda.to_device(self.sat)
-        self.d_ops = cuda.to_device(self.ops)
-        self.d_timing = cuda.to_device(self.timing)
-        self.d_tdata = cuda.to_device(self.tdata)
-        self.d_cdata = cuda.to_device(self.cdata)
-
-        self._block_dim = (32, 16)
-
-    def get_line_delay(self, line, polarity):
-        return self.d_timing[line, 0, polarity]
-
-    def set_line_delay(self, line, polarity, delay):
-        self.d_timing[line, 0, polarity] = delay
-
-    def assign(self, vectors, time=0.0, offset=0):
-        assert (offset % 8) == 0
-        byte_offset = offset // 8
-        assert byte_offset < vectors.bits.shape[-1]
-        pdim = min(vectors.bits.shape[-1] - byte_offset, self.tdata.shape[-1])
-
-        self.tdata[..., 0:pdim] = vectors.bits[..., byte_offset:pdim + byte_offset]
-        if vectors.vdim == 1:
-            self.tdata[:, 1, 0:pdim] = ~self.tdata[:, 1, 0:pdim]
-            self.tdata[:, 2, 0:pdim] = 0
-        cuda.to_device(self.tdata, to=self.d_tdata)
-
-        grid_dim = self._grid_dim(self.sims, len(self.interface))
-        assign_kernel[grid_dim, self._block_dim](self.d_state, self.d_sat, self.ppi_offset,
-                                                 len(self.interface), self.d_tdata, time)
-
-    def _grid_dim(self, x, y):
-        gx = math.ceil(x / self._block_dim[0])
-        gy = math.ceil(y / self._block_dim[1])
-        return gx, gy
-
-    def propagate(self, sims=None, sd=0.0, seed=1):
-        if sims is None:
-            sims = self.sims
-        else:
-            sims = min(sims, self.sims)
-        for op_start, op_stop in zip(self.level_starts, self.level_stops):
-            grid_dim = self._grid_dim(sims, op_stop - op_start)
-            wave_kernel[grid_dim, self._block_dim](self.d_ops, op_start, op_stop, self.d_state, self.sat, int(0),
-                                                   sims, self.d_timing, sd, seed)
-        cuda.synchronize()
-        self.lst_eat_valid = False
-
-    def wave(self, line, vector):
-        if line < 0:
-            return None
-        mem, wcap, _ = self.sat[line]
-        if mem < 0:
-            return None
-        return self.d_state[mem:mem + wcap, vector]
-    
-    def capture(self, time=TMAX, sd=0, seed=1, cdata=None, offset=0):
-        grid_dim = self._grid_dim(self.sims, len(self.interface))
-        capture_kernel[grid_dim, self._block_dim](self.d_state, self.d_sat, self.ppo_offset,
-                                                  self.d_cdata, time, sd * math.sqrt(2), seed)
-        self.cdata[...] = self.d_cdata
-        if cdata is not None:
-            assert offset < cdata.shape[1]
-            cap_dim = min(cdata.shape[1] - offset, self.sims)
-            cdata[:, offset:cap_dim + offset] = self.cdata[:, 0:cap_dim]
-        self.lst_eat_valid = True
-        return self.cdata
-
-    def reassign(self, time=0.0):
-        grid_dim = self._grid_dim(self.sims, len(self.interface))
-        reassign_kernel[grid_dim, self._block_dim](self.d_state, self.d_sat, self.ppi_offset, self.ppo_offset,
-                                                   self.d_cdata, time)
-        cuda.synchronize()
-        
-    def wavecaps(self):
-        gx = math.ceil(len(self.circuit.lines) / 512)
-        wavecaps_kernel[gx, 512](self.d_state, self.d_sat, self.sims)
-        self.sat[...] = self.d_sat
-        return self.sat[..., 2]
-
-
-@cuda.jit()
-def wavecaps_kernel(state, sat, sims):
-    idx = cuda.grid(1)
-    if idx >= len(sat): return
-    
-    lidx, lcap, _ = sat[idx]
-    if lidx < 0: return
-    
-    wcap = 0
-    for sidx in range(sims):
-        for tidx in range(lcap):
-            t = state[lidx + tidx, sidx]
-            if tidx > wcap:
-                wcap = tidx
-            if t >= TMAX: break
-
-    sat[idx, 2] = wcap + 1
-    
-    
-@cuda.jit()
-def reassign_kernel(state, sat, ppi_offset, ppo_offset, cdata, ppi_time):
-    vector, y = cuda.grid(2)
-    if vector >= state.shape[-1]: return
-    if ppo_offset + y >= len(sat): return
-
-    ppo, ppo_cap, _ = sat[ppo_offset + y]
-    ppi, ppi_cap, _ = sat[ppi_offset + y]
-    if ppo < 0: return
-    if ppi < 0: return
-
-    ppo_val = int(cdata[y, vector, 1])
-    ppi_val = int(0)
-    for tidx in range(ppi_cap):
-        t = state[ppi + tidx, vector]
-        if t >= TMAX: break
-        ppi_val ^= 1
-    
-    # make new waveform at PPI
-    toggle = 0
-    if ppi_val:
-        state[ppi + toggle, vector] = TMIN
-        toggle += 1
-    if ppi_val != ppo_val:
-        state[ppi + toggle, vector] = ppi_time
-        toggle += 1
-    state[ppi + toggle, vector] = TMAX
-
-
-@cuda.jit()
-def capture_kernel(state, sat, ppo_offset, cdata, time, s_sqrt2, seed):
-    x, y = cuda.grid(2)
-    if ppo_offset + y >= len(sat): return
-    line, tdim, _ = sat[ppo_offset + y]
-    if line < 0: return
-    if x >= state.shape[-1]: return
-    vector = x
-    m = 0.5
-    acc = 0.0
-    eat = TMAX
-    lst = TMIN
-    tog = 0
-    ovl = 0
-    val = int(0)
-    final = int(0)
-    for tidx in range(tdim):
-        t = state[line + tidx, vector]
-        if t >= TMAX:
-            if t == TMAX_OVL:
-                ovl = 1
-            break
-        m = -m
-        final ^= 1
-        if t < time:
-            val ^= 1
-        if t <= TMIN: continue
-        if s_sqrt2 > 0:
-            acc += m * (1 + math.erf((t - time) / s_sqrt2))
-        eat = min(eat, t)
-        lst = max(lst, t)
-        tog += 1
-    if s_sqrt2 > 0:
-        if m < 0:
-            acc += 1
-        if acc >= 0.99:
-            val = 1
-        elif acc > 0.01:
-            seed = (seed << 4) + (vector << 20) + (y << 1)
-            seed = int(0xDEECE66D) * seed + 0xB
-            seed = int(0xDEECE66D) * seed + 0xB
-            rnd = float((seed >> 8) & 0xffffff) / float(1 << 24)
-            val = rnd < acc
-        else:
-            val = 0
-    else:
-        acc = val
-        
-    cdata[y, vector, 0] = acc
-    cdata[y, vector, 1] = val
-    cdata[y, vector, 2] = final
-    cdata[y, vector, 3] = (val != final)
-    cdata[y, vector, 4] = eat
-    cdata[y, vector, 5] = lst
-    cdata[y, vector, 6] = ovl
-
-
-@cuda.jit()
-def assign_kernel(state, sat, ppi_offset, intf_len, tdata, time):
-    x, y = cuda.grid(2)
-    if y >= intf_len: return
-    line = sat[ppi_offset + y, 0]
-    if line < 0: return
-    sdim = state.shape[-1]
-    if x >= sdim: return
-    vector = x
-    a0 = tdata[y, 0, vector // 8]
-    a1 = tdata[y, 1, vector // 8]
-    a2 = tdata[y, 2, vector // 8]
-    m = np.uint8(1 << (7 - (vector % 8)))
-    toggle = 0
-    if a0 & m:
-        state[line + toggle, x] = TMIN
-        toggle += 1
-    if (a2 & m) and ((a0 & m) == (a1 & m)):
-        state[line + toggle, x] = time
-        toggle += 1
-    state[line + toggle, x] = TMAX
-
-
-@cuda.jit(device=True)
-def rand_gauss(seed, sd):
-    clamp = 0.5
-    if sd <= 0.0:
-        return 1.0
-    while True:
-        x = -6.0
-        for i in range(12):
-            seed = int(0xDEECE66D) * seed + 0xB
-            x += float((seed >> 8) & 0xffffff) / float(1 << 24)
-        x *= sd
-        if abs(x) <= clamp:
-            break
-    return x + 1.0
-
-
-@cuda.jit()
-def wave_kernel(ops, op_start, op_stop, state, sat, st_start, st_stop, line_times, sd, seed):
-    x, y = cuda.grid(2)
-    st_idx = st_start + x
-    op_idx = op_start + y
-    if st_idx >= st_stop: return
-    if op_idx >= op_stop: return
-    lut = ops[op_idx, 0]
-    z_idx = ops[op_idx, 1]
-    a_idx = ops[op_idx, 2]
-    b_idx = ops[op_idx, 3]
-    overflows = int(0)
-
-    _seed = (seed << 4) + (z_idx << 20) + (st_idx << 1)
-
-    a_mem = sat[a_idx, 0]
-    b_mem = sat[b_idx, 0]
-    z_mem, z_cap, _ = sat[z_idx]
-
-    a_cur = int(0)
-    b_cur = int(0)
-    z_cur = lut & 1
-    if z_cur == 1:
-        state[z_mem, st_idx] = TMIN
-
-    a = state[a_mem, st_idx] + line_times[a_idx, 0, z_cur] * rand_gauss(_seed ^ a_mem ^ z_cur, sd)
-    b = state[b_mem, st_idx] + line_times[b_idx, 0, z_cur] * rand_gauss(_seed ^ b_mem ^ z_cur, sd)
-
-    previous_t = TMIN
-
-    current_t = min(a, b)
-    inputs = int(0)
-
-    while current_t < TMAX:
-        z_val = z_cur & 1
-        if b < a:
-            b_cur += 1
-            b = state[b_mem + b_cur, st_idx]
-            b += line_times[b_idx, 0, z_val ^ 1] * rand_gauss(_seed ^ b_mem ^ z_val ^ 1, sd)
-            thresh = line_times[b_idx, 1, z_val] * rand_gauss(_seed ^ b_mem ^ z_val, sd)
-            inputs ^= 2
-            next_t = b
-        else:
-            a_cur += 1
-            a = state[a_mem + a_cur, st_idx]
-            a += line_times[a_idx, 0, z_val ^ 1] * rand_gauss(_seed ^ a_mem ^ z_val ^ 1, sd)
-            thresh = line_times[a_idx, 1, z_val] * rand_gauss(_seed ^ a_mem ^ z_val, sd)
-            inputs ^= 1
-            next_t = a
-
-        if (z_cur & 1) != ((lut >> inputs) & 1):
-            # we generate a toggle in z_mem, if:
-            #   ( it is the first toggle in z_mem OR
-            #   following toggle is earlier OR
-            #   pulse is wide enough ) AND enough space in z_mem.
-            if z_cur == 0 or next_t < current_t or (current_t - previous_t) > thresh:
-                if z_cur < (z_cap - 1):
-                    state[z_mem + z_cur, st_idx] = current_t
-                    previous_t = current_t
-                    z_cur += 1
-                else:
-                    overflows += 1
-                    previous_t = state[z_mem + z_cur - 1, st_idx]
-                    z_cur -= 1
-            else:
-                z_cur -= 1
-                if z_cur > 0:
-                    previous_t = state[z_mem + z_cur - 1, st_idx]
-                else:
-                    previous_t = TMIN
-        current_t = min(a, b)
-        
-    if overflows > 0:
-        state[z_mem + z_cur, st_idx] = TMAX_OVL
-    else:
-        state[z_mem + z_cur, st_idx] = a if a > b else b  # propagate overflow flags by storing biggest TMAX from input

tests/Makefile

@@ -0,0 +1,2 @@
+all:
+	abc -F map_to_minimal.abc

tests/all_kyupy_simprims.minimal.v

@@ -0,0 +1,91 @@
+// Benchmark "all_kyupy_primitives" written by ABC on Thu Nov  6 13:13:21 2025
+
+module all_kyupy_primitives ( 
+    i0, i1, i2, i3,
+    \o[0] , \o[1] , \o[2] , \o[3] , \o[4] , \o[5] , \o[6] , \o[7] , \o[8] ,
+    \o[9] , \o[10] , \o[11] , \o[12] , \o[13] , \o[14] , \o[15] , \o[16] ,
+    \o[17] , \o[18] , \o[19] , \o[20] , \o[21] , \o[22] , \o[23] , \o[24] ,
+    \o[25] , \o[26] , \o[27] , \o[28] , \o[29] , \o[30] , \o[31] , \o[32]   );
+  input  i0, i1, i2, i3;
+  output \o[0] , \o[1] , \o[2] , \o[3] , \o[4] , \o[5] , \o[6] , \o[7] ,
+    \o[8] , \o[9] , \o[10] , \o[11] , \o[12] , \o[13] , \o[14] , \o[15] ,
+    \o[16] , \o[17] , \o[18] , \o[19] , \o[20] , \o[21] , \o[22] , \o[23] ,
+    \o[24] , \o[25] , \o[26] , \o[27] , \o[28] , \o[29] , \o[30] , \o[31] ,
+    \o[32] ;
+  wire new_n45, new_n48, new_n51, new_n54, new_n55, new_n56, new_n57,
+    new_n60, new_n61, new_n62, new_n63, new_n64, new_n65, new_n66, new_n67,
+    new_n68, new_n69, new_n72, new_n73, new_n74, new_n75, new_n76, new_n77,
+    new_n78, new_n79, new_n80, new_n81, new_n86, new_n87, new_n91, new_n92,
+    new_n96, new_n99, new_n102, new_n103, new_n104;
+  INV1 g00(.i0(i1), .o(\o[1] ));
+  AND2 g01(.i0(i1), .i1(i0), .o(\o[2] ));
+  AND2 g02(.i0(\o[2] ), .i1(i2), .o(\o[3] ));
+  AND2 g03(.i0(\o[3] ), .i1(i3), .o(\o[4] ));
+  INV1 g04(.i0(\o[2] ), .o(\o[5] ));
+  INV1 g05(.i0(\o[3] ), .o(\o[6] ));
+  INV1 g06(.i0(\o[4] ), .o(\o[7] ));
+  INV1 g07(.i0(i0), .o(new_n45));
+  AND2 g08(.i0(\o[1] ), .i1(new_n45), .o(\o[11] ));
+  INV1 g09(.i0(\o[11] ), .o(\o[8] ));
+  INV1 g10(.i0(i2), .o(new_n48));
+  AND2 g11(.i0(\o[11] ), .i1(new_n48), .o(\o[12] ));
+  INV1 g12(.i0(\o[12] ), .o(\o[9] ));
+  INV1 g13(.i0(i3), .o(new_n51));
+  AND2 g14(.i0(\o[12] ), .i1(new_n51), .o(\o[13] ));
+  INV1 g15(.i0(\o[13] ), .o(\o[10] ));
+  AND2 g16(.i0(\o[1] ), .i1(i0), .o(new_n54));
+  INV1 g17(.i0(new_n54), .o(new_n55));
+  AND2 g18(.i0(i1), .i1(new_n45), .o(new_n56));
+  INV1 g19(.i0(new_n56), .o(new_n57));
+  AND2 g20(.i0(new_n57), .i1(new_n55), .o(\o[17] ));
+  INV1 g21(.i0(\o[17] ), .o(\o[14] ));
+  AND2 g22(.i0(i2), .i1(i1), .o(new_n60));
+  INV1 g23(.i0(new_n60), .o(new_n61));
+  AND2 g24(.i0(new_n48), .i1(\o[1] ), .o(new_n62));
+  INV1 g25(.i0(new_n62), .o(new_n63));
+  AND2 g26(.i0(new_n63), .i1(new_n61), .o(new_n64));
+  INV1 g27(.i0(new_n64), .o(new_n65));
+  AND2 g28(.i0(new_n65), .i1(i0), .o(new_n66));
+  INV1 g29(.i0(new_n66), .o(new_n67));
+  AND2 g30(.i0(new_n64), .i1(new_n45), .o(new_n68));
+  INV1 g31(.i0(new_n68), .o(new_n69));
+  AND2 g32(.i0(new_n69), .i1(new_n67), .o(\o[18] ));
+  INV1 g33(.i0(\o[18] ), .o(\o[15] ));
+  AND2 g34(.i0(i3), .i1(new_n48), .o(new_n72));
+  INV1 g35(.i0(new_n72), .o(new_n73));
+  AND2 g36(.i0(new_n51), .i1(i2), .o(new_n74));
+  INV1 g37(.i0(new_n74), .o(new_n75));
+  AND2 g38(.i0(new_n75), .i1(new_n73), .o(new_n76));
+  INV1 g39(.i0(new_n76), .o(new_n77));
+  AND2 g40(.i0(new_n77), .i1(\o[17] ), .o(new_n78));
+  INV1 g41(.i0(new_n78), .o(new_n79));
+  AND2 g42(.i0(new_n76), .i1(\o[14] ), .o(new_n80));
+  INV1 g43(.i0(new_n80), .o(new_n81));
+  AND2 g44(.i0(new_n81), .i1(new_n79), .o(\o[19] ));
+  INV1 g45(.i0(\o[19] ), .o(\o[16] ));
+  AND2 g46(.i0(\o[5] ), .i1(new_n48), .o(\o[24] ));
+  INV1 g47(.i0(\o[24] ), .o(\o[20] ));
+  AND2 g48(.i0(i3), .i1(i2), .o(new_n86));
+  INV1 g49(.i0(new_n86), .o(new_n87));
+  AND2 g50(.i0(new_n87), .i1(\o[5] ), .o(\o[25] ));
+  INV1 g51(.i0(\o[25] ), .o(\o[21] ));
+  AND2 g52(.i0(\o[8] ), .i1(i2), .o(\o[22] ));
+  AND2 g53(.i0(new_n51), .i1(new_n48), .o(new_n91));
+  INV1 g54(.i0(new_n91), .o(new_n92));
+  AND2 g55(.i0(new_n92), .i1(\o[8] ), .o(\o[23] ));
+  INV1 g56(.i0(\o[22] ), .o(\o[26] ));
+  INV1 g57(.i0(\o[23] ), .o(\o[27] ));
+  AND2 g58(.i0(\o[5] ), .i1(new_n51), .o(new_n96));
+  AND2 g59(.i0(new_n96), .i1(new_n48), .o(\o[30] ));
+  INV1 g60(.i0(\o[30] ), .o(\o[28] ));
+  AND2 g61(.i0(\o[8] ), .i1(i3), .o(new_n99));
+  AND2 g62(.i0(new_n99), .i1(i2), .o(\o[29] ));
+  INV1 g63(.i0(\o[29] ), .o(\o[31] ));
+  AND2 g64(.i0(new_n48), .i1(i0), .o(new_n102));
+  INV1 g65(.i0(new_n102), .o(new_n103));
+  AND2 g66(.i0(new_n103), .i1(new_n61), .o(new_n104));
+  INV1 g67(.i0(new_n104), .o(\o[32] ));
+  BUF1 g68(.i0(i0), .o(\o[0] ));
+endmodule
+
+

tests/all_kyupy_simprims.v

@@ -0,0 +1,53 @@
+module all_kyupy_primitives (i0, i1, i2, i3, o);
+    input i0;
+    input i1;
+    input i2;
+    input i3;
+    output [32:0] o;
+
+    BUF1 buf1_0 (.i0(i0), .o(o[0]));
+    INV1 inv1_0 (.i0(i1), .o(o[1]));
+
+    AND2 and2_0 (.i0(i0), .i1(i1), .o(o[2]));
+    AND3 and3_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[3]));
+    AND4 and4_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[4]));
+
+    NAND2 nand2_0 (.i0(i0), .i1(i1), .o(o[5]));
+    NAND3 nand3_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[6]));
+    NAND4 nand4_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[7]));
+
+    OR2 or2_0 (.i0(i0), .i1(i1), .o(o[8]));
+    OR3 or3_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[9]));
+    OR4 or4_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[10]));
+
+    NOR2 nor2_0 (.i0(i0), .i1(i1), .o(o[11]));
+    NOR3 nor3_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[12]));
+    NOR4 nor4_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[13]));
+
+    XOR2 xor2_0 (.i0(i0), .i1(i1), .o(o[14]));
+    XOR3 xor3_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[15]));
+    XOR4 xor4_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[16]));
+
+    XNOR2 xnor2_0 (.i0(i0), .i1(i1), .o(o[17]));
+    XNOR3 xnor3_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[18]));
+    XNOR4 xnor4_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[19]));
+
+    AO21 ao21_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[20]));
+    AO22 ao22_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[21]));
+    OA21 oa21_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[22]));
+    OA22 oa22_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[23]));
+
+    AOI21 aoi21_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[24]));
+    AOI22 aoi22_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[25]));
+    OAI21 oai21_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[26]));
+    OAI22 oai22_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[27]));
+
+    AO211 ao211_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[28]));
+    OA211 oa211_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[29]));
+
+    AOI211 aoi211_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[30]));
+    OAI211 oai211_0 (.i0(i0), .i1(i1), .i2(i2), .i3(i3), .o(o[31]));
+
+    MUX21 mux21_0 (.i0(i0), .i1(i1), .i2(i2), .o(o[32]));
+
+endmodule

tests/conftest.py

@@ -1,8 +1,56 @@
 import pytest
 
 
-@pytest.fixture
+@pytest.fixture(scope='session')
 def mydir():
     import os
     from pathlib import Path
     return Path(os.path.realpath(os.path.join(os.getcwd(), os.path.dirname(__file__))))
+
+@pytest.fixture(scope='session')
+def b15_2ig_circuit(mydir):
+    from kyupy import verilog
+    from kyupy.techlib import SAED32
+    return verilog.load(mydir / 'b15_2ig.v.gz', branchforks=True, tlib=SAED32)
+
+@pytest.fixture(scope='session')
+def b15_2ig_circuit_resolved(b15_2ig_circuit):
+    from kyupy.techlib import SAED32
+    cr = b15_2ig_circuit.copy()
+    cr.resolve_tlib_cells(SAED32)
+    return cr
+
+@pytest.fixture(scope='session')
+def b15_4ig_circuit(mydir):
+    from kyupy import verilog
+    from kyupy.techlib import SAED32
+    return verilog.load(mydir / 'b15_4ig.v.gz', branchforks=True, tlib=SAED32)
+
+@pytest.fixture(scope='session')
+def b15_4ig_circuit_resolved(b15_4ig_circuit):
+    from kyupy.techlib import SAED32
+    cr = b15_4ig_circuit.copy()
+    cr.resolve_tlib_cells(SAED32)
+    return cr
+
+@pytest.fixture(scope='session')
+def b15_2ig_delays(mydir, b15_2ig_circuit):
+    from kyupy import sdf
+    from kyupy.techlib import SAED32
+    return sdf.load(mydir / 'b15_2ig.sdf.gz').iopaths(b15_2ig_circuit, tlib=SAED32)[1:2]
+
+@pytest.fixture(scope='session')
+def b15_2ig_sa_nf_test_resp(mydir, b15_2ig_circuit_resolved):
+    from kyupy import stil
+    s = stil.load(mydir / 'b15_2ig.sa_nf.stil.gz')
+    tests = s.tests(b15_2ig_circuit_resolved)[:,1:]
+    resp = s.responses(b15_2ig_circuit_resolved)[:,1:]
+    return (tests, resp)
+
+@pytest.fixture(scope='session')
+def b15_4ig_sa_rf_test_resp(mydir, b15_4ig_circuit_resolved):
+    from kyupy import stil
+    s = stil.load(mydir / 'b15_4ig.sa_rf.stil.gz')
+    tests = s.tests(b15_4ig_circuit_resolved)[:,1:]
+    resp = s.responses(b15_4ig_circuit_resolved)[:,1:]
+    return (tests, resp)

tests/gates.sdf

@@ -7,22 +7,49 @@
 (TEMPERATURE 25.00:25.00:25.00)
 (TIMESCALE 1ns)
 (CELL
-  (CELLTYPE "NAND2X1")
+  (CELLTYPE "NAND2_X1")
   (INSTANCE nandgate)
   (DELAY
     (ABSOLUTE
-    (IOPATH IN1 QN (0.099:0.103:0.103) (0.122:0.127:0.127))
-    (IOPATH IN2 QN (0.083:0.086:0.086) (0.100:0.104:0.104))
+    (IOPATH A1 ZN (0.099:0.103:0.103) (0.122:0.127:0.127))
+    (IOPATH A2 ZN (0.083:0.086:0.086) (0.100:0.104:0.104))
     )
   )
 )
 (CELL
-  (CELLTYPE "AND2X1")
+  (CELLTYPE "AND2_X1")
   (INSTANCE andgate)
   (DELAY
     (ABSOLUTE
-    (IOPATH IN1 Q (0.367:0.378:0.378) (0.351:0.377:0.377))
-    (IOPATH IN2 Q (0.366:0.375:0.375) (0.359:0.370:0.370))
+    (IOPATH A1 ZN (0.367:0.378:0.378) (0.351:0.377:0.377))
+    (IOPATH A2 ZN (0.366:0.375:0.375) (0.359:0.370:0.370))
+    )
+  )
+)
+(CELL
+  (CELLTYPE "OAI21_X1")
+  (INSTANCE oai21gate)
+  (DELAY 
+    (ABSOLUTE
+    (IOPATH B1 ZN (0.000:0.025:0.025) (0.000:0.013:0.013))
+    (IOPATH B2 ZN (0.000:0.030:0.030) (0.000:0.016:0.016))
+    ( COND (B1 == 1'b0) && (B2 == 1'b1) (IOPATH A ZN (0.000:0.018:0.018)))
+    (COND (B1 == 1'b1) && (B2 == 1'b0) (IOPATH A ZN (0.000:0.018:0.018) (0.000:0.016:0.016)))
+    (COND (B1 == 1'b1) && (B2 == 1'b1) (IOPATH A ZN (0.000:0.019:0.019) (0.000:0.014:0.014)))
+    )
+  )
+)
+(CELL
+  (CELLTYPE "MUX2_X1")
+  (INSTANCE mux2gate)
+  (DELAY 
+    (ABSOLUTE
+    (COND (B == 1'b0) && (S == 1'b0) (IOPATH A Z (0.000:0.037:0.037) (0.000:0.058:0.058)))
+    (COND (B == 1'b1) && (S == 1'b0) (IOPATH A Z (0.000:0.037:0.037) (0.000:0.058:0.058)))
+    (COND (A == 1'b0) && (S == 1'b1) (IOPATH B Z (0.000:0.035:0.035) (0.000:0.056:0.056)))
+    (COND (A == 1'b1) && (S == 1'b1) (IOPATH B Z (0.000:0.035:0.035) (0.000:0.056:0.056)))
+    (COND (A == 1'b0) && (B == 1'b1) (IOPATH S Z (0.000:0.047:0.047) (0.000:0.073:0.073)))
+    (COND (A == 1'b1) && (B == 1'b0) (IOPATH S Z (0.000:0.072:0.072) (0.000:0.064:0.064)))
     )
   )
 )

tests/gates.v

@@ -1,11 +1,15 @@
-module gates (a, b, o0, o1 );
+module gates (a, b, c, o0, o1, o2, o3 );
 input a;
 input b;
+input c;
 output o0;
 output o1;
+output o2;
+output o3;
 
-AND2X1 andgate (.IN1 ( a ) , .IN2 ( b ) , .Q ( o0 ) ) ;
-NAND2X1 nandgate (.IN1 ( a ) , .IN2 ( b ) , .QN ( o1 ) ) ;
-
+AND2_X1 andgate (.A1 ( a ) , .A2 ( b ) , .ZN ( o0 ) ) ;
+NAND2_X1 nandgate (.A1 ( a ) , .A2 ( b ) , .ZN ( o1 ) ) ;
+OAI21_X1 oai21gate (.B1(a), .B2(b), .A(c), .ZN(o2) ) ;
+MUX2_X1 mux2gate (.A(a), .B(b), .S(c), .Z(o3)) ;
 
 endmodule

tests/kyupy_simprims.genlib

@@ -0,0 +1,34 @@
+# library of all KyuPy simulation primitives defined in kyupy.sim
+GATE BUF1     1  o=i0;                         PIN * NONINV 1 999 1 0 1 0
+GATE INV1     1  o=!i0;                        PIN * INV 1 999 1 0 1 0
+GATE AND2     1  o=i0*i1;                      PIN * NONINV 1 999 1 0 1 0
+GATE AND3     1  o=i0*i1*i2;                   PIN * NONINV 1 999 1 0 1 0
+GATE AND4     1  o=i0*i1*i2*i3;                PIN * NONINV 1 999 1 0 1 0
+GATE NAND2    1  o=!(i0*i1);                   PIN * INV 1 999 1 0 1 0
+GATE NAND3    1  o=!(i0*i1*i2);                PIN * INV 1 999 1 0 1 0
+GATE NAND4    1  o=!(i0*i1*i2*i3);             PIN * INV 1 999 1 0 1 0
+GATE OR2      1  o=i0+i1;                      PIN * NONINV 1 999 1 0 1 0
+GATE OR3      1  o=i0+i1+i2;                   PIN * NONINV 1 999 1 0 1 0
+GATE OR4      1  o=i0+i1+i2+i3;                PIN * NONINV 1 999 1 0 1 0
+GATE NOR2     1  o=!(i0+i1);                   PIN * INV 1 999 1 0 1 0
+GATE NOR3     1  o=!(i0+i1+i2);                PIN * INV 1 999 1 0 1 0
+GATE NOR4     1  o=!(i0+i1+i2+i3);             PIN * INV 1 999 1 0 1 0
+GATE XOR2     1  o=i0^i1;                      PIN * UNKNOWN 1 999 1 0 1 0
+GATE XOR3     1  o=i0^i1^i2;                   PIN * UNKNOWN 1 999 1 0 1 0
+GATE XOR4     1  o=i0^i1^i2^i3;                PIN * UNKNOWN 1 999 1 0 1 0
+GATE XNOR2    1  o=!(i0^i1);                   PIN * UNKNOWN 1 999 1 0 1 0
+GATE XNOR3    1  o=!(i0^i1^i2);                PIN * UNKNOWN 1 999 1 0 1 0
+GATE XNOR4    1  o=!(i0^i1^i2^i3);             PIN * UNKNOWN 1 999 1 0 1 0
+GATE AO21     1  o=(i0*i1)+i2;                 PIN * NONINV 1 999 1 0 1 0
+GATE AO22     1  o=(i0*i1)+(i2*i3);            PIN * NONINV 1 999 1 0 1 0
+GATE OA21     1  o=(i0+i1)*i2;                 PIN * NONINV 1 999 1 0 1 0
+GATE OA22     1  o=(i0+i1)*(i2+i3);            PIN * NONINV 1 999 1 0 1 0
+GATE AOI21    1  o=!( (i0*i1)+i2 );            PIN * INV 1 999 1 0 1 0
+GATE AOI22    1  o=!( (i0*i1)+(i2*i3) );       PIN * INV 1 999 1 0 1 0
+GATE OAI21    1  o=!( (i0+i1)*i2 );            PIN * INV 1 999 1 0 1 0
+GATE OAI22    1  o=!( (i0+i1)*(i2+i3) );       PIN * INV 1 999 1 0 1 0
+GATE AO211    1  o=(i0*i1)+i2+i3;              PIN * NONINV 1 999 1 0 1 0
+GATE OA211    1  o=(i0+i1)*i2*i3;              PIN * NONINV 1 999 1 0 1 0
+GATE AOI211   1  o=!( (i0*i1)+i2+i3 );         PIN * INV 1 999 1 0 1 0
+GATE OAI211   1  o=!( (i0+i1)*i2*i3 );         PIN * INV 1 999 1 0 1 0
+GATE MUX21    1  o=(i0*!i2)+(i1*i2);           PIN * UNKNOWN 1 999 1 0 1 0

tests/map_to_minimal.abc

@@ -0,0 +1,8 @@
+# abc -f map_to_minimal.abc
+
+read kyupy_simprims.genlib
+read -m all_kyupy_simprims.v
+fraig
+read minimal.genlib
+map
+write all_kyupy_simprims.minimal.v

tests/minimal.genlib

@@ -0,0 +1,5 @@
+# A minimal library for generating logically equivalent circuits with
+# minimum number of gate types.
+GATE BUF1     1  o=i0;                         PIN * NONINV 1 999 1 0 1 0
+GATE INV1     1  o=!i0;                        PIN * INV 1 999 1 0 1 0
+GATE AND2     1  o=i0*i1;                      PIN * NONINV 1 999 1 0 1 0

tests/rng_haltonBase2.synth_yosys.v

@@ -0,0 +1,335 @@
+/* Generated by Yosys 0.9 (git sha1 UNKNOWN, gcc 4.8.5 -fPIC -Os) */
+
+(* top =  1  *)
+(* src = "rng_haltonBase2.v:1" *)
+module rng1(clk, reset, o_output);
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  wire [11:0] _00_;
+  wire _01_;
+  wire _02_;
+  wire _03_;
+  wire _04_;
+  wire _05_;
+  wire _06_;
+  wire _07_;
+  wire _08_;
+  wire _09_;
+  wire _10_;
+  wire _11_;
+  wire _12_;
+  wire _13_;
+  wire _14_;
+  wire _15_;
+  wire _16_;
+  wire _17_;
+  wire _18_;
+  wire _19_;
+  wire _20_;
+  wire _21_;
+  wire _22_;
+  wire _23_;
+  wire _24_;
+  wire _25_;
+  wire _26_;
+  wire _27_;
+  wire _28_;
+  wire _29_;
+  wire _30_;
+  wire _31_;
+  wire _32_;
+  wire _33_;
+  wire _34_;
+  (* src = "rng_haltonBase2.v:2" *)
+  input clk;
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:12" *)
+  wire \halton.clk ;
+  (* init = 12'h000 *)
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:17" *)
+  wire [11:0] \halton.counter ;
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:14" *)
+  wire [11:0] \halton.o_output ;
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:13" *)
+  wire \halton.reset ;
+  (* src = "rng_haltonBase2.v:4" *)
+  output [11:0] o_output;
+  (* src = "rng_haltonBase2.v:3" *)
+  input reset;
+  AND2X1 _35_ (
+    .IN1(\halton.counter [1]),
+    .IN2(\halton.counter [0]),
+    .Q(_01_)
+  );
+  NOR2X0 _36_ (
+    .IN1(\halton.counter [1]),
+    .IN2(\halton.counter [0]),
+    .QN(_02_)
+  );
+  NOR3X0 _37_ (
+    .IN1(reset),
+    .IN2(_01_),
+    .IN3(_02_),
+    .QN(_00_[1])
+  );
+  AND2X1 _38_ (
+    .IN1(\halton.counter [2]),
+    .IN2(_01_),
+    .Q(_03_)
+  );
+  NOR2X0 _39_ (
+    .IN1(\halton.counter [2]),
+    .IN2(_01_),
+    .QN(_04_)
+  );
+  NOR3X0 _40_ (
+    .IN1(reset),
+    .IN2(_03_),
+    .IN3(_04_),
+    .QN(_00_[2])
+  );
+  AND4X1 _41_ (
+    .IN1(\halton.counter [1]),
+    .IN2(\halton.counter [0]),
+    .IN3(\halton.counter [2]),
+    .IN4(\halton.counter [3]),
+    .Q(_05_)
+  );
+  NOR2X0 _42_ (
+    .IN1(\halton.counter [3]),
+    .IN2(_03_),
+    .QN(_06_)
+  );
+  NOR3X0 _43_ (
+    .IN1(reset),
+    .IN2(_05_),
+    .IN3(_06_),
+    .QN(_00_[3])
+  );
+  AND2X1 _44_ (
+    .IN1(\halton.counter [4]),
+    .IN2(_05_),
+    .Q(_07_)
+  );
+  NOR2X0 _45_ (
+    .IN1(\halton.counter [4]),
+    .IN2(_05_),
+    .QN(_08_)
+  );
+  NOR3X0 _46_ (
+    .IN1(reset),
+    .IN2(_07_),
+    .IN3(_08_),
+    .QN(_00_[4])
+  );
+  AND2X1 _47_ (
+    .IN1(\halton.counter [5]),
+    .IN2(_07_),
+    .Q(_09_)
+  );
+  NOR2X0 _48_ (
+    .IN1(\halton.counter [5]),
+    .IN2(_07_),
+    .QN(_10_)
+  );
+  NOR3X0 _49_ (
+    .IN1(reset),
+    .IN2(_09_),
+    .IN3(_10_),
+    .QN(_00_[5])
+  );
+  AND4X1 _50_ (
+    .IN1(\halton.counter [4]),
+    .IN2(\halton.counter [5]),
+    .IN3(\halton.counter [6]),
+    .IN4(_05_),
+    .Q(_11_)
+  );
+  NOR2X0 _51_ (
+    .IN1(\halton.counter [6]),
+    .IN2(_09_),
+    .QN(_12_)
+  );
+  NOR3X0 _52_ (
+    .IN1(reset),
+    .IN2(_11_),
+    .IN3(_12_),
+    .QN(_00_[6])
+  );
+  AND2X1 _53_ (
+    .IN1(\halton.counter [7]),
+    .IN2(_11_),
+    .Q(_13_)
+  );
+  NOR2X0 _54_ (
+    .IN1(\halton.counter [7]),
+    .IN2(_11_),
+    .QN(_14_)
+  );
+  NOR3X0 _55_ (
+    .IN1(reset),
+    .IN2(_13_),
+    .IN3(_14_),
+    .QN(_00_[7])
+  );
+  AND3X1 _56_ (
+    .IN1(\halton.counter [7]),
+    .IN2(\halton.counter [8]),
+    .IN3(_11_),
+    .Q(_15_)
+  );
+  NOR2X0 _57_ (
+    .IN1(\halton.counter [8]),
+    .IN2(_13_),
+    .QN(_16_)
+  );
+  NOR3X0 _58_ (
+    .IN1(reset),
+    .IN2(_15_),
+    .IN3(_16_),
+    .QN(_00_[8])
+  );
+  AND4X1 _59_ (
+    .IN1(\halton.counter [7]),
+    .IN2(\halton.counter [8]),
+    .IN3(\halton.counter [9]),
+    .IN4(_11_),
+    .Q(_17_)
+  );
+  NOR2X0 _60_ (
+    .IN1(\halton.counter [9]),
+    .IN2(_15_),
+    .QN(_18_)
+  );
+  NOR3X0 _61_ (
+    .IN1(reset),
+    .IN2(_17_),
+    .IN3(_18_),
+    .QN(_00_[9])
+  );
+  AND2X1 _62_ (
+    .IN1(\halton.counter [10]),
+    .IN2(_17_),
+    .Q(_19_)
+  );
+  NOR2X0 _63_ (
+    .IN1(\halton.counter [10]),
+    .IN2(_17_),
+    .QN(_20_)
+  );
+  NOR3X0 _64_ (
+    .IN1(reset),
+    .IN2(_19_),
+    .IN3(_20_),
+    .QN(_00_[10])
+  );
+  AND3X1 _65_ (
+    .IN1(\halton.counter [10]),
+    .IN2(\halton.counter [11]),
+    .IN3(_17_),
+    .Q(_21_)
+  );
+  AOI21X1 _66_ (
+    .IN1(\halton.counter [10]),
+    .IN2(_17_),
+    .IN3(\halton.counter [11]),
+    .QN(_22_)
+  );
+  NOR3X0 _67_ (
+    .IN1(reset),
+    .IN2(_21_),
+    .IN3(_22_),
+    .QN(_00_[11])
+  );
+  NOR2X0 _68_ (
+    .IN1(reset),
+    .IN2(\halton.counter [0]),
+    .QN(_00_[0])
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _69_ (
+    .CLK(clk),
+    .D(_00_[0]),
+    .Q(\halton.counter [0]),
+    .QN(_23_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _70_ (
+    .CLK(clk),
+    .D(_00_[1]),
+    .Q(\halton.counter [1]),
+    .QN(_24_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _71_ (
+    .CLK(clk),
+    .D(_00_[2]),
+    .Q(\halton.counter [2]),
+    .QN(_25_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _72_ (
+    .CLK(clk),
+    .D(_00_[3]),
+    .Q(\halton.counter [3]),
+    .QN(_26_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _73_ (
+    .CLK(clk),
+    .D(_00_[4]),
+    .Q(\halton.counter [4]),
+    .QN(_27_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _74_ (
+    .CLK(clk),
+    .D(_00_[5]),
+    .Q(\halton.counter [5]),
+    .QN(_28_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _75_ (
+    .CLK(clk),
+    .D(_00_[6]),
+    .Q(\halton.counter [6]),
+    .QN(_29_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _76_ (
+    .CLK(clk),
+    .D(_00_[7]),
+    .Q(\halton.counter [7]),
+    .QN(_30_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _77_ (
+    .CLK(clk),
+    .D(_00_[8]),
+    .Q(\halton.counter [8]),
+    .QN(_31_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _78_ (
+    .CLK(clk),
+    .D(_00_[9]),
+    .Q(\halton.counter [9]),
+    .QN(_32_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _79_ (
+    .CLK(clk),
+    .D(_00_[10]),
+    .Q(\halton.counter [10]),
+    .QN(_33_)
+  );
+  (* src = "rng_haltonBase2.v:7|rng_haltonBase2.v:19" *)
+  DFFX1 _80_ (
+    .CLK(clk),
+    .D(_00_[11]),
+    .Q(\halton.counter [11]),
+    .QN(_34_)
+  );
+  assign \halton.clk  = clk;
+  assign \halton.o_output  = { \halton.counter [0], \halton.counter [1], \halton.counter [2], \halton.counter [3], \halton.counter [4], \halton.counter [5], \halton.counter [6], \halton.counter [7], \halton.counter [8], \halton.counter [9], \halton.counter [10], \halton.counter [11] };
+  assign \halton.reset  = reset;
+  assign o_output = { \halton.counter [0], \halton.counter [1], \halton.counter [2], \halton.counter [3], \halton.counter [4], \halton.counter [5], \halton.counter [6], \halton.counter [7], \halton.counter [8], \halton.counter [9], \halton.counter [10], \halton.counter [11] };
+endmodule

tests/test_atalanta.py

@@ -0,0 +1,29 @@
+import numpy as np
+
+from kyupy import atalanta, logic
+
+
+def test_parse():
+    text = '\n'.join([
+        '* Test pattern file',
+        '1: 0011',
+        '2: 1100',
+        '1010',
+    ])
+    tf = atalanta.parse(text)
+
+    # second-to-last axis = inputs, last axis = patterns
+    assert tf.patterns.shape == (4, 3)
+    assert tf.patterns.dtype == np.uint8
+    assert set(np.unique(tf.patterns)) <= {logic.ZERO, logic.ONE}
+
+    # patterns are stored along the last axis
+    assert list(tf.patterns[:, 0]) == [logic.ZERO, logic.ZERO, logic.ONE, logic.ONE]
+    assert list(tf.patterns[:, 1]) == [logic.ONE, logic.ONE, logic.ZERO, logic.ZERO]
+    assert list(tf.patterns[:, 2]) == [logic.ONE, logic.ZERO, logic.ONE, logic.ZERO]
+
+
+def test_empty():
+    tf = atalanta.parse('* only comments\n* nothing else\n')
+    assert tf.patterns.shape == (0, 0)
+    assert tf.patterns.dtype == np.uint8

tests/test_bench.py

@@ -4,12 +4,12 @@ from kyupy import bench
 def test_b01(mydir):
     with open(mydir / 'b01.bench', 'r') as f:
         c = bench.parse(f.read())
-        assert 92 == len(c.nodes)
-    c = bench.parse(mydir / 'b01.bench')
-    assert 92 == len(c.nodes)
+        assert len(c.nodes) == 92
+    c = bench.load(mydir / 'b01.bench')
+    assert len(c.nodes) == 92
 
 
 def test_simple():
     c = bench.parse('input(a, b) output(z) z=and(a,b)')
     assert len(c.nodes) == 4
-    assert len(c.interface) == 3
+    assert len(c.io_nodes) == 3

tests/test_circuit.py

@@ -1,4 +1,73 @@
-from kyupy.circuit import Circuit, Node, Line
+import pickle
+
+from kyupy.circuit import GrowingList, Circuit, Node, Line
+from kyupy import verilog, bench
+from kyupy.techlib import KYUPY, SAED32
+
+def test_growing_list():
+    gl = GrowingList()
+    assert gl.free_idx == 0
+    gl[0] = 1
+    assert gl.free_idx == 1
+    gl[2] = 1
+    assert gl.free_idx == 1
+    gl[0] = None
+    assert gl.free_idx == 0
+    gl[0] = 1
+    assert gl.free_idx == 1
+    gl[1] = 1
+    assert gl.free_idx == 3
+    gl.append(1)
+    assert gl.free_idx == 4
+    gl[2] = None
+    assert gl.free_idx == 2
+    gl[2] = 1
+    gl[1] = None
+    assert gl.free_idx == 1
+
+def test_lines():
+    c = Circuit()
+    n1 = Node(c, 'n1')
+    n2 = Node(c, 'n2')
+    line = Line(c, n1, n2)
+
+    assert line.driver == n1
+    assert line.reader == n2
+    assert line.driver_pin == 0
+    assert line.reader_pin == 0
+    assert n1.outs[0] == line
+    assert n2.ins[0] == line
+
+    line2 = Line(c, n1, (n2, 2))
+
+    assert line2.driver == n1
+    assert line2.reader == n2
+    assert line2.driver_pin == 1
+    assert line2.reader_pin == 2
+    assert n1.outs[0] == line
+    assert n1.outs[1] == line2
+    assert n2.ins[1] is None
+    assert n2.ins[2] == line2
+
+    line3 = Line(c, n1, n2)
+
+    assert line3.driver_pin == 2
+    assert line3.reader_pin == 1
+    assert n1.outs[2] == line3
+    assert n2.ins[1] == line3
+    assert n2.ins[2] == line2
+
+    assert len(c.lines) == 3
+
+    line3.remove()
+
+    assert len(c.lines) == 2
+    assert c.lines[0].index == 0
+    assert c.lines[1].index == 1
+
+    assert len(n1.outs) == 2
+    assert n2.ins[1] is None
+    assert n2.ins[2] == line2
 
 
 def test_circuit():
@@ -10,9 +79,9 @@ def test_circuit():
     assert 'in1' in c.cells
     assert 'and1' not in c.cells
 
-    c.interface[0] = in1
-    c.interface[1] = in2
-    c.interface[2] = out1
+    c.io_nodes[0] = in1
+    c.io_nodes[1] = in2
+    c.io_nodes[2] = out1
 
     and1 = Node(c, 'and1', kind='and')
     Line(c, in1, and1)
@@ -54,3 +123,39 @@ def test_circuit():
 
     for n in c.topological_order():
         repr(n)
+
+
+def test_pickle(mydir):
+    c = verilog.load(mydir / 'b15_4ig.v.gz', tlib=SAED32)
+    assert c is not None
+    cs = pickle.dumps(c)
+    assert cs is not None
+    c2 = pickle.loads(cs)
+    assert len(c.nodes) == len(c2.nodes)
+
+def test_equivalency():
+    c1 = bench.parse('input(a) output(z) z=not(a)')
+    c2 = bench.parse('input(a) output(z) z=not(a)')
+    assert c1 != c2
+    assert c1.lines[0] != c2.lines[0]
+    assert c1.nodes[0] != c2.nodes[0]
+
+
+def test_substitute():
+    c = bench.parse('input(i1, i2, i3, i4, i5) output(o1) aoi=AOI221(i1, i2, i3, i4, i5) o1=not(aoi)')
+    assert len(c.cells) == 2
+    assert len(c.io_nodes) == 6
+    aoi221_impl = bench.parse('input(in1, in2, in3, in4, in5) output(q) a1=and(in1, in2) a2=and(in3, in4) q=or(a1, a2, in5)')
+    assert len(aoi221_impl.cells) == 3
+    assert len(aoi221_impl.io_nodes) == 6
+    c.substitute(c.cells['aoi'], aoi221_impl)
+    assert len(c.cells) == 4
+    assert len(c.io_nodes) == 6
+
+
+def test_resolve(mydir):
+    c = verilog.load(mydir / 'b15_4ig.v.gz', tlib=SAED32)
+    s_names = [n.name for n in c.s_nodes(SAED32)]
+    c.resolve_tlib_cells(SAED32)
+    s_names_prim = [n.name for n in c.s_nodes(KYUPY)]
+    assert s_names == s_names_prim, 'resolve_tlib_cells does not preserve names or order of s_nodes'

tests/test_logic.py

@@ -0,0 +1,75 @@
+import numpy as np
+import kyupy.logic as lg
+from kyupy.logic import mvarray, bparray, bp_to_mv, mv_to_bp
+
+
+def assert_equal_shape_and_contents(actual, desired):
+    desired = np.array(desired, dtype=np.uint8)
+    assert actual.shape == desired.shape
+    np.testing.assert_allclose(actual, desired)
+
+
+def test_mvarray_single_vector():
+    assert_equal_shape_and_contents(mvarray(1, 0, 1), [lg.ONE, lg.ZERO, lg.ONE])
+    assert_equal_shape_and_contents(mvarray([1, 0, 1]), [lg.ONE, lg.ZERO, lg.ONE])
+    assert_equal_shape_and_contents(mvarray('10X-RFPN'), [lg.ONE, lg.ZERO, lg.UNKNOWN, lg.UNASSIGNED, lg.RISE, lg.FALL, lg.PPULSE, lg.NPULSE])
+    assert_equal_shape_and_contents(mvarray(['1']), [lg.ONE])
+    assert_equal_shape_and_contents(mvarray('1'), [lg.ONE])
+
+
+def test_mvarray_multi_vector():
+    assert_equal_shape_and_contents(mvarray([0, 0], [0, 1], [1, 0], [1, 1]), [[lg.ZERO, lg.ZERO, lg.ONE, lg.ONE], [lg.ZERO, lg.ONE, lg.ZERO, lg.ONE]])
+    assert_equal_shape_and_contents(mvarray('10X', '--1'), [[lg.ONE, lg.UNASSIGNED], [lg.ZERO, lg.UNASSIGNED], [lg.UNKNOWN, lg.ONE]])
+
+
+def test_mv_ops():
+    x1_8v = mvarray('00000000XXXXXXXX--------11111111PPPPPPPPRRRRRRRRFFFFFFFFNNNNNNNN')
+    x2_8v = mvarray('0X-1PRFN'*8)
+
+    assert_equal_shape_and_contents(lg.mv_not(x1_8v), mvarray('11111111XXXXXXXXXXXXXXXX00000000NNNNNNNNFFFFFFFFRRRRRRRRPPPPPPPP'))
+    assert_equal_shape_and_contents(lg.mv_or(x1_8v, x2_8v), mvarray('0XX1PRFNXXX1XXXXXXX1XXXX11111111PXX1PRFNRXX1RRNNFXX1FNFNNXX1NNNN'))
+    assert_equal_shape_and_contents(lg.mv_and(x1_8v, x2_8v), mvarray('000000000XXXXXXX0XXXXXXX0XX1PRFN0XXPPPPP0XXRPRPR0XXFPPFF0XXNPRFN'))
+    assert_equal_shape_and_contents(lg.mv_xor(x1_8v, x2_8v), mvarray('0XX1PRFNXXXXXXXXXXXXXXXX1XX0NFRPPXXNPRFNRXXFRPNFFXXRFNPRNXXPNFRP'))
+
+    # TODO
+    #assert_equal_shape_and_contents(lg.mv_transition(x1_8v, x2_8v), mvarray('0XXR PRFNXXXXXXXXXXXXXXXX1XX0NFRPPXXNPRFNRXXFRPNFFXXRFNPRNXXPNFRP'))
+
+    x30_8v = mvarray('0000000000000000000000000000000000000000000000000000000000000000')
+    x31_8v = mvarray('1111111111111111111111111111111111111111111111111111111111111111')
+
+    assert_equal_shape_and_contents(lg.mv_latch(x1_8v, x2_8v, x30_8v), mvarray('0XX000000XXXXXXX0XXXXXXX0XX10R110XX000000XXR0R0R0XXF001F0XX10R11'))
+    assert_equal_shape_and_contents(lg.mv_latch(x1_8v, x2_8v, x31_8v), mvarray('1XX01F001XXXXXXX1XXXXXXX1XX111111XX01F001XXR110R1XXF1F1F1XX11111'))
+
+
+def test_bparray():
+
+    bpa = bparray('0X-1PRFN')
+    assert bpa.shape == (8, 3, 1)
+
+    bpa = bparray('0X-1PRFN-')
+    assert bpa.shape == (9, 3, 1)
+
+    bpa = bparray('000', '001', '010', '011', '100', '101', '110', '111')
+    assert bpa.shape == (3, 3, 1)
+
+    bpa = bparray('000', '001', '010', '011', '100', '101', '110', '111', 'RFX')
+    assert bpa.shape == (3, 3, 2)
+
+    assert_equal_shape_and_contents(bp_to_mv(bparray('0X-1PRFN'))[:,0], mvarray('0X-1PRFN'))
+    assert_equal_shape_and_contents(bparray('0X-1PRFN'), mv_to_bp(mvarray('0X-1PRFN')))
+
+    x1_8v = bparray('00000000XXXXXXXX--------11111111PPPPPPPPRRRRRRRRFFFFFFFFNNNNNNNN')
+    x2_8v = bparray('0X-1PRFN'*8)
+
+    out_8v = np.empty((64, 3, 1), dtype=np.uint8)
+
+    assert_equal_shape_and_contents(bp_to_mv(lg.bp8v_buf(out_8v, x1_8v))[:,0], mvarray('00000000XXXXXXXXXXXXXXXX11111111PPPPPPPPRRRRRRRRFFFFFFFFNNNNNNNN'))
+    assert_equal_shape_and_contents(bp_to_mv(lg.bp8v_or(out_8v, x1_8v, x2_8v))[:,0], mvarray('0XX1PRFNXXX1XXXXXXX1XXXX11111111PXX1PRFNRXX1RRNNFXX1FNFNNXX1NNNN'))
+    assert_equal_shape_and_contents(bp_to_mv(lg.bp8v_and(out_8v, x1_8v, x2_8v))[:,0], mvarray('000000000XXXXXXX0XXXXXXX0XX1PRFN0XXPPPPP0XXRPRPR0XXFPPFF0XXNPRFN'))
+    assert_equal_shape_and_contents(bp_to_mv(lg.bp8v_xor(out_8v, x1_8v, x2_8v))[:,0], mvarray('0XX1PRFNXXXXXXXXXXXXXXXX1XX0NFRPPXXNPRFNRXXFRPNFFXXRFNPRNXXPNFRP'))
+
+    x30_8v = bparray('0000000000000000000000000000000000000000000000000000000000000000')
+    x31_8v = bparray('1111111111111111111111111111111111111111111111111111111111111111')
+
+    assert_equal_shape_and_contents(bp_to_mv(lg.bp8v_latch(out_8v, x1_8v, x2_8v, x30_8v))[:,0], mvarray('0XX000000XXXXXXX0XXXXXXX0XX10R110XX000000XXR0R0R0XXF001F0XX10R11'))
+    assert_equal_shape_and_contents(bp_to_mv(lg.bp8v_latch(out_8v, x1_8v, x2_8v, x31_8v))[:,0], mvarray('1XX01F001XXXXXXX1XXXXXXX1XX111111XX01F001XXR110R1XXF1F1F1XX11111'))

tests/test_logic_sim.py

@@ -1,161 +1,396 @@
-from kyupy.logic_sim import LogicSim
-from kyupy import bench
-from kyupy.packed_vectors import PackedVectors
+import numpy as np
+
+from kyupy.logic_sim import LogicSim, LogicSim2V, LogicSim4V, LogicSim6V
+from kyupy import bench, logic, sim, verilog
+from kyupy.logic import mv_str, mvarray, bparray, bp_to_mv, mv_to_bp
+from kyupy.techlib import SAED90, KYUPY
+
+def test_dangling():
+    c = verilog.parse('''
+        module test(i, o);
+            input i; output o; wire dangling;
+            INVX0 dff1 (.INP(i),.ZN(o));
+            assign dangling=o;
+        endmodule
+    ''', tlib=SAED90)
+    c.resolve_tlib_cells(SAED90)
+    sim = LogicSim2V(c)
+    p = sim.allocate()
+    p[...] = 0
+    sim.simulate(p)
+    assert p[1,0] == logic.ONE
+
+def test_used_output():
+    c = bench.parse('input(i0) output(o0) output(o1) o0=BUF1(i0) o1=BUF(o0)')
+    sim = LogicSim2V(c)
+    expect = mvarray('000', '111')
+    actual = sim.simulate(expect.copy())
+    np.testing.assert_array_equal(
+        expect[c.io_locs('o0')],
+        actual[c.io_locs('o0')])
+    np.testing.assert_array_equal(
+        expect[c.io_locs('o1')],
+        actual[c.io_locs('o1')])
+
+def assert_equal_shape_and_contents(actual, desired):
+    desired = np.array(desired, dtype=np.uint8)
+    assert actual.shape == desired.shape
+    np.testing.assert_allclose(actual, desired)
+
+def test_LogicSim2V_minimal():
+    c = bench.parse('input(i0, i1) output(o0, o1, o2) o0=AND2(i0,i1) o1=INV1(i0) o2=BUF1(i0)')
+    sim = LogicSim2V(c)
+    expect = mvarray('00010',
+                     '01010',
+                     '10001',
+                     '11101')
+    actual = sim.simulate(expect.copy())
+    np.testing.assert_array_equal(
+        expect[c.io_locs('o')],
+        actual[c.io_locs('o')])
+
+def test_LogicSim4V_minimal():
+    c = bench.parse('input(i0, i1) output(o0, o1, o2) o0=AND2(i0,i1) o1=INV1(i0) o2=BUF1(i0)')
+    sim = LogicSim4V(c, 16)
+    expect = mvarray('00010', '01010', '0-010', '0X010',
+                     '10001', '11101', '1-X01', '1XX01',
+                     '-00XX', '-1XXX', '--XXX', '-XXXX',
+                     'X00XX', 'X1XXX', 'X-XXX', 'XXXXX',)
+    actual = sim.simulate(expect.copy())
+    np.testing.assert_array_equal(
+        expect[c.io_locs('o')],
+        actual[c.io_locs('o')])
+
+def test_LogicSim6V_minimal():
+    c = bench.parse('input(i0, i1) output(o0, o1, o2) o0=AND2(i0,i1) o1=INV1(i0) o2=BUF1(i0)')
+    sim = LogicSim6V(c, 36)
+    expect = mvarray('00010', '01010', '0R010', '0F010', '0P010', '0N010',
+                     '10001', '11101', '1RR01', '1FF01', '1PP01', '1NN01',
+                     'R00FR', 'R1RFR', 'RRRFR', 'RFPFR', 'RPPFR', 'RNRFR',
+                     'F00RF', 'F1FRF', 'FRPRF', 'FFFRF', 'FPPRF', 'FNFRF',
+                     'P00NP', 'P1PNP', 'PRPNP', 'PFPNP', 'PPPNP', 'PNPNP',
+                     'N00PN', 'N1NPN', 'NRRPN', 'NFFPN', 'NPPPN', 'NNNPN')
+    actual = sim.simulate(expect.copy())
+    np.testing.assert_array_equal(
+        expect[c.io_locs('o')],
+        actual[c.io_locs('o')])
+
+def test_LogicSim2V_simprims(mydir):
+    c1 = verilog.load(mydir / 'all_kyupy_simprims.v')
+    c2 = verilog.load(mydir / 'all_kyupy_simprims.minimal.v')
+    pi_count = sum([len(n.ins)==0 for n in c1.io_nodes])
+    sim1 = LogicSim2V(c1, sims=2**pi_count)
+    sim2 = LogicSim2V(c2, sims=2**pi_count)
+    tests1 = sim1.allocate()
+    tests1[sim1.pippi_s_locs] = np.fromfunction(lambda x, y: logic.ONE * ((2**x & y) != 0), (pi_count, sim1.sims), dtype=int)  # exhaustive test
+    tests2 = tests1.copy()
+    sim1.simulate(tests1)
+    sim2.simulate(tests2)
+    np.testing.assert_array_equal(
+        tests1[c1.io_locs('o')],
+        tests2[c2.io_locs('o')])
+
+def test_LogicSim4V_simprims(mydir):
+    c1 = verilog.load(mydir / 'all_kyupy_simprims.v')
+    c2 = verilog.load(mydir / 'all_kyupy_simprims.minimal.v')
+    pi_count = sum([len(n.ins)==0 for n in c1.io_nodes])
+    sim1 = LogicSim4V(c1, sims=4**pi_count)
+    sim2 = LogicSim4V(c2, sims=4**pi_count)
+    tests1 = sim1.allocate()
+    tests1[sim1.pippi_s_locs] = np.fromfunction(lambda x, y: (y//(4**x))%4, (pi_count, sim1.sims), dtype=int)  # exhaustive test
+    tests2 = tests1.copy()
+    sim1.simulate(tests1)
+    sim2.simulate(tests2)
+    np.testing.assert_array_equal(
+        tests1[c1.io_locs('o')],
+        tests2[c2.io_locs('o')])
+
+def test_LogicSim6V_simprims(mydir):
+    c1 = verilog.load(mydir / 'all_kyupy_simprims.v')
+    c2 = verilog.load(mydir / 'all_kyupy_simprims.minimal.v')
+    pi_count = sum([len(n.ins)==0 for n in c1.io_nodes])
+    sim1 = LogicSim6V(c1, sims=6**pi_count)
+    sim2 = LogicSim6V(c2, sims=6**pi_count)
+    tests1 = sim1.allocate()
+    tests1[sim1.pippi_s_locs] = logic.mvarray('01RFPN')[np.fromfunction(lambda x, y: (y//(6**x))%6, (pi_count, sim1.sims), dtype=int)]  # exhaustive test
+    tests2 = tests1.copy()
+    sim1.simulate(tests1)
+    sim2.simulate(tests2)
+    for loc1, loc2 in zip(c1.io_locs('o'), c2.io_locs('o')):
+        n = c1.s_nodes(KYUPY)[loc1]
+        if (tests1[loc1] != tests2[loc2]).any():
+            print(f'Mismatch at output {n}')
+    np.testing.assert_array_equal(
+        tests1[c1.io_locs('o')],
+        tests2[c2.io_locs('o')])
+
+def test_2v():
+    c = bench.parse(f'''
+        input(i3, i2, i1, i0)
+        output({",".join([f"o{i:02d}" for i in range(33)])})
+        o00=BUF1(i0)
+        o01=INV1(i0)
+        o02=AND2(i0,i1)
+        o03=AND3(i0,i1,i2)
+        o04=AND4(i0,i1,i2,i3)
+        o05=NAND2(i0,i1)
+        o06=NAND3(i0,i1,i2)
+        o07=NAND4(i0,i1,i2,i3)
+        o08=OR2(i0,i1)
+        o09=OR3(i0,i1,i2)
+        o10=OR4(i0,i1,i2,i3)
+        o11=NOR2(i0,i1)
+        o12=NOR3(i0,i1,i2)
+        o13=NOR4(i0,i1,i2,i3)
+        o14=XOR2(i0,i1)
+        o15=XOR3(i0,i1,i2)
+        o16=XOR4(i0,i1,i2,i3)
+        o17=XNOR2(i0,i1)
+        o18=XNOR3(i0,i1,i2)
+        o19=XNOR4(i0,i1,i2,i3)
+        o20=AO21(i0,i1,i2)
+        o21=OA21(i0,i1,i2)
+        o22=AO22(i0,i1,i2,i3)
+        o23=OA22(i0,i1,i2,i3)
+        o24=AOI21(i0,i1,i2)
+        o25=OAI21(i0,i1,i2)
+        o26=AOI22(i0,i1,i2,i3)
+        o27=OAI22(i0,i1,i2,i3)
+        o28=AO211(i0,i1,i2,i3)
+        o29=OA211(i0,i1,i2,i3)
+        o30=AOI211(i0,i1,i2,i3)
+        o31=OAI211(i0,i1,i2,i3)
+        o32=MUX21(i0,i1,i2)
+    ''')
+    s = LogicSim2V(c, 16)
+    s.s_assign[...] = logic.mvarray([f'{i:04b}'+('-'*(s.s_len-4)) for i in range(16)])
+    s.s_to_c()
+    s.c_prop()
+    s.c_to_s()
+    for res, exp in zip(logic.packbits(s.s_result[4:], dtype=np.uint32), [
+            sim.BUF1, sim.INV1,
+            sim.AND2, sim.AND3, sim.AND4,
+            sim.NAND2, sim.NAND3, sim.NAND4,
+            sim.OR2, sim.OR3, sim.OR4,
+            sim.NOR2, sim.NOR3, sim.NOR4,
+            sim.XOR2, sim.XOR3, sim.XOR4,
+            sim.XNOR2, sim.XNOR3, sim.XNOR4,
+            sim.AO21, sim.OA21,
+            sim.AO22, sim.OA22,
+            sim.AOI21, sim.OAI21,
+            sim.AOI22, sim.OAI22,
+            sim.AO211, sim.OA211,
+            sim.AOI211, sim.OAI211,
+            sim.MUX21
+        ]):
+        assert res == exp, f'Mismatch for SimPrim {sim.names[exp]} res={bin(res)} exp={bin(exp)}'
 
 
-def test_vd1():
+def test_4v():
     c = bench.parse('input(x, y) output(a, o, n) a=and(x,y) o=or(x,y) n=not(x)')
-    s = LogicSim(c, 4)
-    assert len(s.interface) == 5
-    p = PackedVectors(4, len(s.interface))
-    p[0] = '00000'
-    p[1] = '01000'
-    p[2] = '10000'
-    p[3] = '11000'
-    s.assign(p)
-    s.propagate()
-    s.capture(p)
-    assert p[0] == '00001'
-    assert p[1] == '01011'
-    assert p[2] == '10010'
-    assert p[3] == '11110'
+    s = LogicSim4V(c, 16)
+    assert s.s_len == 5
+    s.s_assign[...] = mvarray(
+        '00---', '01---', '0----', '0X---',
+        '10---', '11---', '1----', '1X---',
+        '-0---', '-1---', '-----', '-X---',
+        'X0---', 'X1---', 'X----', 'XX---')
+    s.s_to_c()
+    s.c_prop()
+    s.c_to_s()
+    assert_equal_shape_and_contents(s.s_result, mvarray(
+        '--001', '--011', '--0X1', '--0X1',
+        '--010', '--110', '--X10', '--X10',
+        '--0XX', '--X1X', '--XXX', '--XXX',
+        '--0XX', '--X1X', '--XXX', '--XXX'))
+
+def test_4v_fault():
+    c = bench.parse('input(x, y) output(a) a=and(x,y)')
+    s = LogicSim4V(c, 16)
+    assert s.s_len == 3
+    s.s_assign[...] = mvarray(
+        '00-', '01-', '0--', '0X-',
+        '10-', '11-', '1--', '1X-',
+        '-0-', '-1-', '---', '-X-',
+        'X0-', 'X1-', 'X--', 'XX-')
+    s.s_to_c()
+    s.c_prop()
+    s.c_to_s()
+    assert_equal_shape_and_contents(s.s_result, mvarray(
+        '--0', '--0', '--0', '--0',
+        '--0', '--1', '--X', '--X',
+        '--0', '--X', '--X', '--X',
+        '--0', '--X', '--X', '--X'))
+    fault_line = s.circuit.cells['a'].ins[0]
+    s.s_to_c()
+    s.c_prop(fault_line=fault_line.index, fault_model=1)
+    s.c_to_s()
+    assert_equal_shape_and_contents(s.s_result, mvarray(
+        '--0', '--1', '--X', '--X',
+        '--0', '--1', '--X', '--X',
+        '--0', '--1', '--X', '--X',
+        '--0', '--1', '--X', '--X'))
+    s.s_to_c()
+    s.c_prop(fault_line=fault_line.index, fault_model=0)
+    s.c_to_s()
+    assert_equal_shape_and_contents(s.s_result, mvarray(
+        '--0', '--0', '--0', '--0',
+        '--0', '--0', '--0', '--0',
+        '--0', '--0', '--0', '--0',
+        '--0', '--0', '--0', '--0'))
+
+def test_6v():
+    c = bench.parse('input(x, y) output(a, o, n, xo, no) a=AND2(x,y) o=OR2(x,y) n=INV1(x) xo=XOR2(x,y) no=NOR2(x,y)')
+    s = LogicSim6V(c, 36)
+    assert s.s_len == 7
+    mva = mvarray(
+        '0000101', '0101110', '0R0R1RF', '0F0F1FR', '0P0P1PN', '0N0N1NP',
+        '1001010', '1111000', '1RR10F0', '1FF10R0', '1PP10N0', '1NN10P0',
+        'R00RFRF', 'R1R1FF0', 'RRRRFPF', 'RFPNFNP', 'RPPRFRF', 'RNRNFFP',
+        'F00FRFR', 'F1F1RR0', 'FRPNRNP', 'FFFFRPR', 'FPPFRFR', 'FNFNRRP',
+        'P00PNPN', 'P1P1NN0', 'PRPRNRF', 'PFPFNFR', 'PPPPNPN', 'PNPNNNP',
+        'N00NPNP', 'N1N1PP0', 'NRRNPFP', 'NFFNPRP', 'NPPNPNP', 'NNNNPPP')
+    tests = np.copy(mva)
+    tests[2:] = logic.ZERO
+    s.s[0] = tests
+    s.s_to_c()
+    s.c_prop()
+    s.c_to_s()
+    resp = s.s[1].copy()
+
+    exp_resp = np.copy(mva)
+    exp_resp[:2] = logic.UNASSIGNED
+    np.testing.assert_allclose(resp, exp_resp)
 
 
-def test_vd2():
-    c = bench.parse('input(x, y) output(a, o, n) a=and(x,y) o=or(x,y) n=not(x)')
-    s = LogicSim(c, 16, 2)
-    assert len(s.interface) == 5
-    p = PackedVectors(16, len(s.interface), 2)
-    p[0] = '00000'
-    p[1] = '01000'
-    p[2] = '0-000'
-    p[3] = '0X000'
-    p[4] = '10000'
-    p[5] = '11000'
-    p[6] = '1-000'
-    p[7] = '1X000'
-    p[8] = '-0000'
-    p[9] = '-1000'
-    p[10] = '--000'
-    p[11] = '-X000'
-    p[12] = 'X0000'
-    p[13] = 'X1000'
-    p[14] = 'X-000'
-    p[15] = 'XX000'
-    s.assign(p)
-    s.propagate()
-    s.capture(p)
-    assert p[0] == '00001'
-    assert p[1] == '01011'
-    assert p[2] == '0-0X1'
-    assert p[3] == '0X0X1'
-    assert p[4] == '10010'
-    assert p[5] == '11110'
-    assert p[6] == '1-X10'
-    assert p[7] == '1XX10'
-    assert p[8] == '-00XX'
-    assert p[9] == '-1X1X'
-    assert p[10] == '--XXX'
-    assert p[11] == '-XXXX'
-    assert p[12] == 'X00XX'
-    assert p[13] == 'X1X1X'
-    assert p[14] == 'X-XXX'
-    assert p[15] == 'XXXXX'
-
-    
-def test_vd3():
+def xtest_8v():
     c = bench.parse('input(x, y) output(a, o, n, xo) a=and(x,y) o=or(x,y) n=not(x) xo=xor(x,y)')
-    s = LogicSim(c, 64, 3)
-    assert len(s.interface) == 6
-    p = PackedVectors(64, len(s.interface), 3)
-    p[0] = '000010'
-    p[1] = '010111'
-    p[2] = '0-0X1X'
-    p[3] = '0X0X1X'
-    p[4] = '0R0R1R'
-    p[5] = '0F0F1F'
-    p[6] = '0P0P1P'
-    p[7] = '0N0N1N'
-    p[8] = '100101'
-    p[9] = '111100'
-    p[10] = '1-X10X'
-    p[11] = '1XX10X'
-    p[12] = '1RR10F'
-    p[13] = '1FF10R'
-    p[14] = '1PP10N'
-    p[15] = '1NN10P'
-    p[16] = '-00XXX'
-    p[17] = '-1X1XX'
-    p[18] = '--XXXX'
-    p[19] = '-XXXXX'
-    p[20] = '-RXXXX'
-    p[21] = '-FXXXX'
-    p[22] = '-PXXXX'
-    p[23] = '-NXXXX'
-    p[24] = 'X00XXX'
-    p[25] = 'X1X1XX'
-    p[26] = 'X-XXXX'
-    p[27] = 'XXXXXX'
-    p[28] = 'XRXXXX'
-    p[29] = 'XFXXXX'
-    p[30] = 'XPXXXX'
-    p[31] = 'XNXXXX'
-    p[32] = 'R00RFR'
-    p[33] = 'R1R1FF'
-    p[34] = 'R-XXFX'
-    p[35] = 'RXXXFX'
-    p[36] = 'RRRRFP'
-    p[37] = 'RFPNFN'
-    p[38] = 'RPPRFR'
-    p[39] = 'RNRNFF'
-    p[40] = 'F00FRF'
-    p[41] = 'F1F1RR'
-    p[42] = 'F-XXRX'
-    p[43] = 'FXXXRX'
-    p[44] = 'FRPNRN'
-    p[45] = 'FFFFRP'
-    p[46] = 'FPPFRF'
-    p[47] = 'FNFNRR'
-    p[48] = 'P00PNP'
-    p[49] = 'P1P1NN'
-    p[50] = 'P-XXNX'
-    p[51] = 'PXXXNX'
-    p[52] = 'PRPRNR'
-    p[53] = 'PFPFNF'
-    p[54] = 'PPPPNP'
-    p[55] = 'PNPNNN'
-    p[56] = 'N00NPN'
-    p[57] = 'N1N1PP'
-    p[58] = 'N-XXPX'
-    p[59] = 'NXXXPX'
-    p[60] = 'NRRNPF'
-    p[61] = 'NFFNPR'
-    p[62] = 'NPPNPN'
-    p[63] = 'NNNNPP'
-    expect = p.copy()
-    s.assign(p)
-    s.propagate()
-    s.capture(p)
-    for i in range(64):
-        assert p[i] == expect[i]
-        
+    s = LogicSim(c, 64, m=8)
+    assert s.s_len == 6
+    mva = mvarray(
+        '000010', '010111', '0-0X1X', '0X0X1X', '0R0R1R', '0F0F1F', '0P0P1P', '0N0N1N',
+        '100101', '111100', '1-X10X', '1XX10X', '1RR10F', '1FF10R', '1PP10N', '1NN10P',
+        '-00XXX', '-1X1XX', '--XXXX', '-XXXXX', '-RXXXX', '-FXXXX', '-PXXXX', '-NXXXX',
+        'X00XXX', 'X1X1XX', 'X-XXXX', 'XXXXXX', 'XRXXXX', 'XFXXXX', 'XPXXXX', 'XNXXXX',
+        'R00RFR', 'R1R1FF', 'R-XXFX', 'RXXXFX', 'RRRRFP', 'RFPNFN', 'RPPRFR', 'RNRNFF',
+        'F00FRF', 'F1F1RR', 'F-XXRX', 'FXXXRX', 'FRPNRN', 'FFFFRP', 'FPPFRF', 'FNFNRR',
+        'P00PNP', 'P1P1NN', 'P-XXNX', 'PXXXNX', 'PRPRNR', 'PFPFNF', 'PPPPNP', 'PNPNNN',
+        'N00NPN', 'N1N1PP', 'N-XXPX', 'NXXXPX', 'NRRNPF', 'NFFNPR', 'NPPNPN', 'NNNNPP')
+    tests = np.copy(mva)
+    tests[2:] = logic.UNASSIGNED
+    bpa = mv_to_bp(tests)
+    s.s[0] = bpa
+    s.s_to_c()
+    s.c_prop()
+    s.c_to_s()
+    resp = bp_to_mv(s.s[1])
 
-def test_b01(mydir):
-    c = bench.parse(mydir / 'b01.bench')
+    exp_resp = np.copy(mva)
+    exp_resp[:2] = logic.UNASSIGNED
+    np.testing.assert_allclose(resp, exp_resp)
 
-    # 2-valued
-    s = LogicSim(c, 8)
-    assert len(s.interface) == 9
-    t = PackedVectors(8, len(s.interface))
-    t.randomize()
-    s.assign(t)
-    s.propagate()
-    s.capture(t)
 
-    # 8-valued
-    s = LogicSim(c, 8, 3)
-    t = PackedVectors(8, len(s.interface), 3)
-    t.randomize()
-    s.assign(t)
-    s.propagate()
-    s.capture(t)
+def xtest_loop():
+    c = bench.parse('q=DFF(d) d=NOT(q)')
+    s = LogicSim(c, 4, m=8)
+    assert s.s_len == 1
+    mva = mvarray([['0'], ['1'], ['R'], ['F']])
+
+    # TODO
+    # s.assign(BPArray(mva))
+    # s.propagate()
+    # resp_bp = BPArray((len(s.interface), s.sims))
+    # s.capture(resp_bp)
+    # resp = MVArray(resp_bp)
+
+    # assert resp[0] == '1'
+    # assert resp[1] == '0'
+    # assert resp[2] == 'F'
+    # assert resp[3] == 'R'
+
+    # resp_bp = s.cycle(resp_bp)
+    # resp = MVArray(resp_bp)
+
+    # assert resp[0] == '0'
+    # assert resp[1] == '1'
+    # assert resp[2] == 'R'
+    # assert resp[3] == 'F'
+
+
+def xtest_latch():
+    c = bench.parse('input(d, t) output(q) q=LATCH(d, t)')
+    s = LogicSim(c, 8, m=8)
+    assert s.s_len == 4
+    mva = mvarray('00-0', '00-1', '01-0', '01-1', '10-0', '10-1', '11-0', '11-1')
+    exp = mvarray('0000', '0011', '0100', '0100', '1000', '1011', '1111', '1111')
+
+    # TODO
+    # resp = MVArray(s.cycle(BPArray(mva)))
+
+    # for i in range(len(mva)):
+    #     assert resp[i] == exp[i]
+
+
+def sim_and_compare(c, test_resp, m=8):
+    tests, resp = test_resp
+    lsim = LogicSim(c, m=m, sims=tests.shape[1])
+    lsim.s[0] = logic.mv_to_bp(tests)
+    lsim.s_to_c()
+    lsim.c_prop()
+    lsim.c_to_s()
+    resp_sim = logic.bp_to_mv(lsim.s[1])[:,:tests.shape[1]]
+    idxs, pats = np.nonzero(((resp == logic.ONE) & (resp_sim != logic.ONE)) | ((resp == logic.ZERO) & (resp_sim != logic.ZERO)))
+    for i, (idx, pat) in enumerate(zip(idxs, pats)):
+        if i >= 10:
+            print(f'...')
+            break
+        print(f'mismatch pattern:{pat} ppio:{idx} exp:{logic.mv_str(resp[idx,pat])} act:{logic.mv_str(resp_sim[idx,pat])}')
+    assert len(idxs) == 0
+
+def sim_and_compare_cls(c, test_resp, SimCls):
+    tests, resp = test_resp
+    lsim = SimCls(c, sims=tests.shape[1])
+    lsim.s_assign[...] = tests
+    lsim.s_to_c()
+    lsim.c_prop()
+    lsim.c_to_s()
+    resp_sim = lsim.s_result
+    idxs, pats = np.nonzero(((resp == logic.ONE) & (resp_sim != logic.ONE)) | ((resp == logic.ZERO) & (resp_sim != logic.ZERO)))
+    for i, (idx, pat) in enumerate(zip(idxs, pats)):
+        if i >= 10:
+            print(f'...')
+            break
+        print(f'mismatch pattern:{pat} ppio:{idx} exp:{logic.mv_str(resp[idx,pat])} act:{logic.mv_str(resp_sim[idx,pat])}')
+    assert len(idxs) == 0
+
+
+def test_b15_2ig_sa_2v(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp):
+    sim_and_compare_cls(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp, LogicSim2V)
+
+
+def test_b15_2ig_sa_4v(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp):
+    sim_and_compare_cls(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp, LogicSim4V)
+
+
+def test_b15_2ig_sa_6v(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp):
+    sim_and_compare_cls(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp, LogicSim6V)
+
+
+def xtest_b15_2ig_sa_8v(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp):
+    sim_and_compare(b15_2ig_circuit_resolved, b15_2ig_sa_nf_test_resp, m=8)
+
+
+def test_b15_4ig_sa_2v(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp):
+    sim_and_compare_cls(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp, LogicSim2V)
+
+
+def test_b15_4ig_sa_4v(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp):
+    sim_and_compare_cls(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp, LogicSim4V)
+
+
+def test_b15_4ig_sa_6v(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp):
+    sim_and_compare_cls(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp, LogicSim6V)
+
+
+def xtest_b15_4ig_sa_8v(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp):
+    sim_and_compare(b15_4ig_circuit_resolved, b15_4ig_sa_rf_test_resp, m=8)

tests/test_packed_vectors.py

@@ -1,88 +0,0 @@
-from kyupy.packed_vectors import PackedVectors
-
-
-def test_basic():
-    ba = PackedVectors(8, 1, 1)
-    assert '0\n0\n0\n0\n0\n0\n0\n0' == str(ba)
-    ba.set_value(0, 0, 1)
-    ba.set_value(1, 0, 'H')
-    ba.set_value(2, 0, 'h')
-    ba.set_value(3, 0, True)
-    ba.set_value(4, 0, 0)
-    ba.set_value(5, 0, 'L')
-    ba.set_value(6, 0, 'l')
-    ba.set_value(7, 0, False)
-    assert '1\n1\n1\n1\n0\n0\n0\n0' == str(ba)
-    ba.set_value(1, 0, '0')
-    ba.set_value(5, 0, '1')
-    assert '1\n0\n1\n1\n0\n1\n0\n0' == str(ba)
-    ba = PackedVectors(8, 1, 2)
-    assert '-\n-\n-\n-\n-\n-\n-\n-' == str(ba)
-    ba.set_value(0, 0, 1)
-    ba.set_value(7, 0, 0)
-    ba.set_value(4, 0, 'X')
-    assert '1\n-\n-\n-\nX\n-\n-\n0' == str(ba)
-    ba.set_value(4, 0, '-')
-    assert '1\n-\n-\n-\n-\n-\n-\n0' == str(ba)
-    ba = PackedVectors(8, 2, 2)
-    assert '--\n--\n--\n--\n--\n--\n--\n--' == str(ba)
-    ba.set_value(0, 0, '1')
-    ba.set_value(7, 1, '0')
-    ba.set_values(1, 'XX')
-    assert '1-\nXX\n--\n--\n--\n--\n--\n-0' == str(ba)
-
-
-def test_8v():
-    ba = PackedVectors(1, 8, 3)
-    assert '--------' == str(ba)
-    ba.set_values(0, r'-x01^v\/')
-    assert r'-X01PNFR' == str(ba)
-    ba.set_values(0, '-XLHPNFR')
-    assert r'-X01PNFR' == str(ba)
-    ba.set_values(0, '-xlhpnfr')
-    assert r'-X01PNFR' == str(ba)
-    p1 = PackedVectors(1, 8, 1)
-    p2 = PackedVectors(1, 8, 1)
-    p1.set_values(0, '01010101')
-    p2.set_values(0, '00110011')
-    p = PackedVectors.from_pair(p1, p2)
-    assert r'0FR10FR1' == str(p)
-    p1 = PackedVectors(1, 8, 2)
-    p2 = PackedVectors(1, 8, 2)
-    p1.set_values(0, '0101-X-X')
-    p2.set_values(0, '00110011')
-    p = PackedVectors.from_pair(p1, p2)
-    assert r'0FR1----' == str(p)
-    p1.set_values(0, '0101-X-X')
-    p2.set_values(0, '-X-X--XX')
-    p = PackedVectors.from_pair(p1, p2)
-    assert r'--------' == str(p)
-
-
-def test_slicing():
-    lv = PackedVectors(3, 2, 1)
-    assert '00\n00\n00' == str(lv)
-    lv.set_value(1, 0, '1')
-    lv.set_value(1, 1, '1')
-    assert '00' == lv[0]
-    assert '11' == lv[1]
-    assert 3 == len(lv)
-    lv2 = lv[1:3]
-    assert 2 == len(lv2)
-    assert '11' == lv2[0]
-    assert '00' == lv2[1]
-
-
-def test_copy():
-    lv1 = PackedVectors(8, 1, 1)
-    lv1.set_values_for_position(0, '01010101')
-    lv2 = PackedVectors(8, 1, 1)
-    lv2.set_values_for_position(0, '00100101')
-    diff = lv1.diff(lv2)
-    lv3 = lv1.copy(selection_mask=diff)
-    assert str(lv3) == '1\n0\n1'
-    lv4 = lv1.copy(selection_mask=~diff)
-    assert str(lv4) == '0\n0\n1\n0\n1'
-    lv5 = lv3 + lv4
-    assert str(lv5) == '1\n0\n1\n0\n0\n1\n0\n1'
-

tests/test_sdf.py

@@ -1,6 +1,8 @@
-from kyupy import sdf, verilog
-from kyupy.saed import pin_index
+import numpy as np
 
+from kyupy import sdf, verilog, bench
+from kyupy.wave_sim import WaveSim, TMAX, TMIN
+from kyupy.techlib import SAED32, NANGATE45
 
 def test_parse():
     test = '''
@@ -17,84 +19,213 @@ def test_parse():
     (TEMPERATURE 25.00:25.00:25.00)
     (TIMESCALE 1ns)
     (CELL
-      (CELLTYPE "b14")
-      (INSTANCE)
-      (DELAY
-        (ABSOLUTE
-        (INTERCONNECT U621/ZN U19246/IN1 (0.000:0.000:0.000))
-        (INTERCONNECT U13292/QN U19246/IN2 (0.001:0.001:0.001))
-        (INTERCONNECT U15050/QN U19247/IN1 (0.000:0.000:0.000))
-        (INTERCONNECT U13293/QN U19247/IN2 (0.000:0.000:0.000) (0.000:0.000:0.000))
+        (CELLTYPE "b14")
+        (INSTANCE)
+        (DELAY
+            (ABSOLUTE
+                (INTERCONNECT U621/ZN U19246/IN1 (0.000:0.000:0.000))
+                (INTERCONNECT U13292/QN U19246/IN2 (0.001:0.001:0.001))
+                (INTERCONNECT U15050/QN U19247/IN1 (0.000:0.000:0.000))
+                (INTERCONNECT U13293/QN U19247/IN2 (0.000:0.000:0.000) (0.000:0.000:0.000))
+            )
         )
-      )
     )
     (CELL
-      (CELLTYPE "INVX2")
-      (INSTANCE U78)
-      (DELAY
-        (ABSOLUTE
-        (IOPATH INP ZN (0.201:0.227:0.227) (0.250:0.271:0.271))
+        (CELLTYPE "INVX2")
+        (INSTANCE U78)
+        (DELAY
+            (ABSOLUTE
+                (IOPATH INP ZN (0.201:0.227:0.227) (0.250:0.271:0.271))
+            )
         )
-      )
     )
     (CELL
-      (CELLTYPE "SDFFARX1")
-      (INSTANCE reg3_reg_1_0)
-      (DELAY
-        (ABSOLUTE
-        (IOPATH (posedge CLK) Q (0.707:0.710:0.710) (0.737:0.740:0.740))
-        (IOPATH (negedge RSTB) Q () (0.909:0.948:0.948))
-        (IOPATH (posedge CLK) QN (0.585:0.589:0.589) (0.545:0.550:0.550))
-        (IOPATH (negedge RSTB) QN (1.546:1.593:1.593) ())
+        (CELLTYPE "SDFFARX1")
+        (INSTANCE reg3_reg_1_0)
+        (DELAY
+            (ABSOLUTE
+                (IOPATH (posedge CLK) Q (0.707:0.710:0.710) (0.737:0.740:0.740))
+                (IOPATH (negedge RSTB) Q () (0.909:0.948:0.948))
+                (IOPATH (posedge CLK) QN (0.585:0.589:0.589) (0.545:0.550:0.550))
+                (IOPATH (negedge RSTB) QN (1.546:1.593:1.593) ())
+            )
         )
-      )
-      (TIMINGCHECK
-        (WIDTH (posedge CLK) (0.284:0.284:0.284))
-        (WIDTH (negedge CLK) (0.642:0.642:0.642))
-        (SETUP (posedge D) (posedge CLK) (0.544:0.553:0.553))
-        (SETUP (negedge D) (posedge CLK) (0.620:0.643:0.643))
-        (HOLD (posedge D) (posedge CLK) (-0.321:-0.331:-0.331))
-        (HOLD (negedge D) (posedge CLK) (-0.196:-0.219:-0.219))
-        (RECOVERY (posedge RSTB) (posedge CLK) (-1.390:-1.455:-1.455))
-        (HOLD (posedge RSTB) (posedge CLK) (1.448:1.509:1.509))
-        (SETUP (posedge SE) (posedge CLK) (0.662:0.670:0.670))
-        (SETUP (negedge SE) (posedge CLK) (0.698:0.702:0.702))
-        (HOLD (posedge SE) (posedge CLK) (-0.435:-0.444:-0.444))
-        (HOLD (negedge SE) (posedge CLK) (-0.291:-0.295:-0.295))
-        (SETUP (posedge SI) (posedge CLK) (0.544:0.544:0.544))
-        (SETUP (negedge SI) (posedge CLK) (0.634:0.688:0.688))
-        (HOLD (posedge SI) (posedge CLK) (-0.317:-0.318:-0.318))
-        (HOLD (negedge SI) (posedge CLK) (-0.198:-0.247:-0.247))
-        (WIDTH (negedge RSTB) (0.345:0.345:0.345))
+        (TIMINGCHECK
+            (WIDTH (posedge CLK) (0.284:0.284:0.284))
+            (WIDTH (negedge CLK) (0.642:0.642:0.642))
+            (SETUP (posedge D) (posedge CLK) (0.544:0.553:0.553))
+            (SETUP (negedge D) (posedge CLK) (0.620:0.643:0.643))
+            (HOLD (posedge D) (posedge CLK) (-0.321:-0.331:-0.331))
+            (HOLD (negedge D) (posedge CLK) (-0.196:-0.219:-0.219))
+            (RECOVERY (posedge RSTB) (posedge CLK) (-1.390:-1.455:-1.455))
+            (HOLD (posedge RSTB) (posedge CLK) (1.448:1.509:1.509))
+            (SETUP (posedge SE) (posedge CLK) (0.662:0.670:0.670))
+            (SETUP (negedge SE) (posedge CLK) (0.698:0.702:0.702))
+            (HOLD (posedge SE) (posedge CLK) (-0.435:-0.444:-0.444))
+            (HOLD (negedge SE) (posedge CLK) (-0.291:-0.295:-0.295))
+            (SETUP (posedge SI) (posedge CLK) (0.544:0.544:0.544))
+            (SETUP (negedge SI) (posedge CLK) (0.634:0.688:0.688))
+            (HOLD (posedge SI) (posedge CLK) (-0.317:-0.318:-0.318))
+            (HOLD (negedge SI) (posedge CLK) (-0.198:-0.247:-0.247))
+            (WIDTH (negedge RSTB) (0.345:0.345:0.345))
     )))
     '''
     df = sdf.parse(test)
     assert df.name == 'test'
-    # print(f'DelayFile(name={df.name}, interconnects={len(df.interconnects)}, iopaths={len(df.iopaths)})')
 
 
-def test_b14(mydir):
-    df = sdf.parse(mydir / 'b14.sdf.gz')
-    assert df.name == 'b14'
+def test_b15(mydir):
+    df = sdf.load(mydir / 'b15_2ig.sdf.gz')
+    assert df.name == 'b15'
 
 
 def test_gates(mydir):
-    c = verilog.parse(mydir / 'gates.v')
-    df = sdf.parse(mydir / 'gates.sdf')
-    lt = df.annotation(c, pin_index, dataset=1)
+    c = verilog.load(mydir / 'gates.v', tlib=NANGATE45)
+    df = sdf.load(mydir / 'gates.sdf')
+    lt = df.iopaths(c, tlib=NANGATE45)[1]
     nand_a = c.cells['nandgate'].ins[0]
     nand_b = c.cells['nandgate'].ins[1]
     and_a = c.cells['andgate'].ins[0]
     and_b = c.cells['andgate'].ins[1]
 
-    assert lt[nand_a.index, 0, 0] == 0.103
-    assert lt[nand_a.index, 0, 1] == 0.127
+    assert lt[nand_a, 0, 0] == 0.103
+    assert lt[nand_a, 0, 1] == 0.127
 
-    assert lt[nand_b.index, 0, 0] == 0.086
-    assert lt[nand_b.index, 0, 1] == 0.104
+    assert lt[nand_b, 0, 0] == 0.086
+    assert lt[nand_b, 0, 1] == 0.104
 
-    assert lt[and_a.index, 0, 0] == 0.378
-    assert lt[and_a.index, 0, 1] == 0.377
+    assert lt[and_a, 0, 0] == 0.378
+    assert lt[and_a, 0, 1] == 0.377
 
-    assert lt[and_b.index, 0, 0] == 0.375
-    assert lt[and_b.index, 0, 1] == 0.370
+    assert lt[and_b, 0, 0] == 0.375
+    assert lt[and_b, 0, 1] == 0.370
+
+
+def test_nand_xor():
+    c = bench.parse("""
+        input(A1,A2)
+        output(lt_1237_U91,lt_1237_U92)
+        lt_1237_U91 = NAND2X0_RVT(A1,A2)
+        lt_1237_U92 = XOR2X1_RVT(A1,A2)
+        """)
+    df = sdf.parse("""
+        (DELAYFILE
+            (CELL
+                (CELLTYPE "NAND2X0_RVT")
+                (INSTANCE lt_1237_U91)
+                (DELAY
+                    (ABSOLUTE
+                        (IOPATH A1 Y (0.018:0.022:0.021) (0.017:0.019:0.019))
+                        (IOPATH A2 Y (0.021:0.024:0.024) (0.018:0.021:0.021))
+                    )
+                )
+            )
+            (CELL
+                (CELLTYPE "XOR2X1_RVT")
+                (INSTANCE lt_1237_U92)
+                (DELAY
+                    (ABSOLUTE
+                        (IOPATH (posedge A1) Y (0.035:0.038:0.038) (0.037:0.062:0.062))
+                        (IOPATH (negedge A1) Y (0.035:0.061:0.061) (0.036:0.040:0.040))
+                        (IOPATH (posedge A2) Y (0.042:0.043:0.043) (0.051:0.064:0.064))
+                        (IOPATH (negedge A2) Y (0.041:0.066:0.066) (0.051:0.053:0.053))
+                    )
+                )
+            )
+        )
+        """)
+    d = df.iopaths(c, tlib=SAED32)[1]
+    c.resolve_tlib_cells(SAED32)
+    sim = WaveSim(c, delays=d, sims=16)
+
+    # input A1
+    sim.s[0,0] = [0,1,0,1] * 4  # initial values  0101010101010101
+    sim.s[1,0] = 0.0            # transition time
+    sim.s[2,0] = [0,0,1,1] * 4  # final values    0011001100110011
+
+    # input A2
+    sim.s[0,1] = ([0]*4 + [1]*4)*2  # initial values  0000111100001111
+    sim.s[1,1] = 0.0                # transition time
+    sim.s[2,1] = [0]*8 + [1]*8      # final values    0000000011111111
+
+    # A1:   0FR10FR10FR10FR1
+    # A2:   0000FFFFRRRR1111
+    # nand: 11111RNR1NFF1RF0
+    # xor:  0FR1FPPRRNPF1RF0
+
+    sim.s_to_c()
+    sim.c_prop()
+    sim.c_to_s()
+
+    eat = sim.s[4,2:]
+    lst = sim.s[5,2:]
+
+    # NAND-gate output
+    assert np.allclose(eat[0], [
+        TMAX, TMAX, TMAX, TMAX, TMAX,
+        0.022,  # FF -> rising Y: min(0.022, 0.024)
+        TMAX,   # RF: pulse filtered
+        0.024,  # falling A2 -> rising Y
+        TMAX,
+        TMAX,   # FR: pulse filtered
+        0.021,  # RR -> falling Y: max(0.019, 0.021)
+        0.021,  # rising A2 -> falling Y
+        TMAX,
+        0.022,  # falling A1 -> rising Y
+        0.019,  # rising A1 -> falling Y
+        TMAX
+    ])
+
+    assert np.allclose(lst[0], [
+        TMIN, TMIN, TMIN, TMIN, TMIN,
+        0.022,  # FF -> rising Y: min(0.022, 0.024)
+        TMIN,   # RF: pulse filtered
+        0.024,  # falling A2 -> rising Y
+        TMIN,
+        TMIN,   # FR: pulse filtered
+        0.021,  # RR -> falling Y: max(0.019, 0.021)
+        0.021,  # rising A2 -> falling Y
+        TMIN,
+        0.022,  # falling A1 -> rising Y
+        0.019,  # rising A1 -> falling Y
+        TMIN
+    ])
+
+    #XOR-gate output
+    assert np.allclose(eat[1], [
+        TMAX,
+        0.040,  # A1:F -> Y:F
+        0.038,  # A1:R -> Y:R
+        TMAX,
+        0.053,  # A2:F -> Y:F
+        TMAX,   # P filtered
+        TMAX,   # P filtered
+        0.066,  # A2:F -> Y:R
+        0.043,  # A2:R -> Y:R
+        TMAX,   # N filtered
+        TMAX,   # P filtered
+        0.064,  # A2:R -> Y:F
+        TMAX,
+        0.061,  # A1:F -> Y:R
+        0.062,  # A1:R -> Y:F
+        TMAX,
+    ])
+
+    assert np.allclose(lst[1], [
+        TMIN,
+        0.040,  # A1:F -> Y:F
+        0.038,  # A1:R -> Y:R
+        TMIN,
+        0.053,  # A2:F -> Y:F
+        TMIN,   # P filtered
+        TMIN,   # P filtered
+        0.066,  # A2:F -> Y:R
+        0.043,  # A2:R -> Y:R
+        TMIN,   # N filtered
+        TMIN,   # P filtered
+        0.064,  # A2:R -> Y:F
+        TMIN,
+        0.061,  # A1:F -> Y:R
+        0.062,  # A1:R -> Y:F
+        TMIN,
+    ])

tests/test_stil.py

@@ -1,9 +1,21 @@
-from kyupy import stil
+from kyupy import stil, verilog
+from kyupy.techlib import SAED32
 
+def test_b15(mydir):
+    b15 = verilog.load(mydir / 'b15_2ig.v.gz', tlib=SAED32)
 
-def test_b14(mydir):
-    s = stil.parse(mydir / 'b14.stuck.stil.gz')
-    assert 10 == len(s.signal_groups)
-    assert 1 == len(s.scan_chains)
-    assert 2163 == len(s.calls)
+    s = stil.load(mydir / 'b15_2ig.sa_nf.stil.gz')
+    assert len(s.signal_groups) == 10
+    assert len(s.scan_chains) == 1
+    assert len(s.calls) == 1357
+    tests = s.tests(b15)
+    resp = s.responses(b15)
+    assert len(tests) > 0
+    assert len(resp) > 0
+
+    s2 = stil.load(mydir / 'b15_2ig.tf_nf.stil.gz')
+    tests = s2.tests_loc(b15, SAED32)
+    resp = s2.responses(b15)
+    assert len(tests) > 0
+    assert len(resp) > 0
 

tests/test_verilog.py

@@ -1,9 +1,45 @@
 from kyupy import verilog
-
+from kyupy.techlib import SAED90, SAED32, NANGATE45
 
 def test_b01(mydir):
     with open(mydir / 'b01.v', 'r') as f:
-        modules = verilog.parse(f.read())
-    assert modules is not None
-    assert verilog.parse(mydir / 'b01.v') is not None
+        c = verilog.parse(f.read(), tlib=SAED90)
+    assert c is not None
+    assert verilog.load(mydir / 'b01.v', tlib=SAED90) is not None
 
+    assert len(c.nodes) == 139
+    assert len(c.lines) == 203
+    stats = c.stats(SAED90)
+    assert stats['input'] == 6
+    assert stats['output'] == 3
+    assert stats['__seq__'] == 5
+
+
+def test_b15(mydir):
+    c = verilog.load(mydir / 'b15_4ig.v.gz', tlib=SAED32)
+    assert len(c.nodes) == 12067
+    assert len(c.lines) == 20731
+    stats = c.stats(SAED32)
+    assert stats['input'] == 40
+    assert stats['output'] == 71
+    assert stats['__seq__'] == 417
+
+
+def test_gates(mydir):
+    c = verilog.load(mydir / 'gates.v', tlib=NANGATE45)
+    assert len(c.nodes) == 18
+    assert len(c.lines) == 21
+    stats = c.stats(NANGATE45)
+    assert stats['input'] == 3
+    assert stats['output'] == 4
+    assert stats['__seq__'] == 0
+
+
+def test_halton2(mydir):
+    c = verilog.load(mydir / 'rng_haltonBase2.synth_yosys.v', tlib=SAED90)
+    assert len(c.nodes) == 146
+    assert len(c.lines) == 210
+    stats = c.stats(SAED90)
+    assert stats['input'] == 2
+    assert stats['output'] == 12
+    assert stats['__seq__'] == 12

tests/test_wave_sim.py

@@ -1,145 +1,203 @@
 import numpy as np
-from kyupy.wave_sim import WaveSim, wave_eval, TMIN, TMAX
-from kyupy.logic_sim import LogicSim
-from kyupy import verilog
-from kyupy import sdf
-from kyupy.saed import pin_index
-from kyupy.packed_vectors import PackedVectors
-from kyupy.wave_sim_cuda import WaveSimCuda
 
+from kyupy.wave_sim import WaveSim, WaveSimCuda, wave_eval_cpu, TMIN, TMAX
+from kyupy.logic_sim import LogicSim6V
+from kyupy import logic, bench, sim
+from kyupy.logic import mvarray
+
+def test_xnor2_delays():
+    op = (sim.XNOR2, 2, 0, 1, 3, 3, -1, 0, 0)
+    #op = (0b0111, 4, 0, 1)
+    c = np.full((4*16, 1), TMAX, dtype=np.float32)  # 4 waveforms of capacity 16
+    c_locs = np.zeros((4,), dtype='int')
+    c_caps = np.zeros((4,), dtype='int')
+    ebuf = np.zeros((4, 1, 2), dtype=np.int32)
+
+    for i in range(4): c_locs[i], c_caps[i] = i*16, 16  # 1:1 mapping
+    delays = np.zeros((1, 4, 2, 2))
+    delays[0, 0, 0, 0] = 0.031  # A rise -> Z rise
+    delays[0, 0, 0, 1] = 0.027  # A rise -> Z fall
+    delays[0, 0, 1, 0] = 0.033  # A fall -> Z rise
+    delays[0, 0, 1, 1] = 0.037  # A fall -> Z fall
+    delays[0, 1, 0, 0] = 0.032  # B rise -> Z rise
+    delays[0, 1, 0, 1] = 0.030  # B rise -> Z fall
+    delays[0, 1, 1, 0] = 0.038  # B fall -> Z rise
+    delays[0, 1, 1, 1] = 0.036  # B fall -> Z fall
+
+    simctl_int = np.asarray([0], dtype=np.int32)
+    simctl_float = np.asarray([1], dtype=np.float32)
+
+    def wave_assert(inputs, output):
+        for i, a in zip(inputs, c.reshape(-1,16)): a[:len(i)] = i
+        wave_eval_cpu(op, c, c_locs, c_caps, ebuf, 0, delays, simctl_int, simctl_float, 0, 0)
+        for i, v in enumerate(output): np.testing.assert_allclose(c.reshape(-1,16)[2,i], v)
+
+    wave_assert([[TMIN,TMAX],[TMIN,TMAX]], [TMIN,TMAX])      # XNOR(1,1) => 1
+    wave_assert([[TMAX,TMAX],[TMIN,TMAX]], [TMAX])           # XNOR(0,1) => 0
+    # using Afall/Zfall for pulse length, bug: was using Arise/Zfall
+    #wave_assert([[0.07, 0.10, TMAX], [0.0, TMAX]], [TMIN, 0.03, 0.101, 0.137, TMAX])
+    wave_assert([[0.07, 0.10, TMAX], [0.0, TMAX]], [TMIN, 0.03, TMAX])
+    wave_assert([[0.06, 0.10, TMAX], [0.0, TMAX]], [TMIN, 0.03, 0.091, 0.137, TMAX])
+
+def test_nand_delays():
+    op = (sim.NAND4, 4, 0, 1, 2, 3, -1, 0, 0)
+    c = np.full((5*16, 1), TMAX, dtype=np.float32)  # 5 waveforms of capacity 16
+    c_locs = np.zeros((5,), dtype='int')
+    c_caps = np.zeros((5,), dtype='int')
+    ebuf = np.zeros((5, 1, 2), dtype=np.int32)
+
+    for i in range(5): c_locs[i], c_caps[i] = i*16, 16  # 1:1 mapping
 
-def test_wave_eval():
     # SDF specifies IOPATH delays with respect to output polarity
     # SDF pulse rejection value is determined by IOPATH causing last transition and polarity of last transition
-    line_times = np.zeros((3, 2, 2))
-    line_times[0, 0, 0] = 0.1  # A -> Z rise delay
-    line_times[0, 0, 1] = 0.2  # A -> Z fall delay
-    line_times[0, 1, 0] = 0.1  # A -> Z negative pulse limit (terminate in rising Z)
-    line_times[0, 1, 1] = 0.2  # A -> Z positive pulse limit
-    line_times[1, 0, 0] = 0.3  # as above for B -> Z
-    line_times[1, 0, 1] = 0.4
-    line_times[1, 1, 0] = 0.3
-    line_times[1, 1, 1] = 0.4
-    
-    state = np.zeros((3*16, 1)) + TMAX  # 3 waveforms of capacity 16
-    state[::16, 0] = 16  # first entry is capacity
-    a = state[0:16, 0]
-    b = state[16:32, 0]
-    z = state[32:, 0]
-    sat = np.zeros((3, 3), dtype='int')
-    sat[0] = 0, 16, 0
-    sat[1] = 16, 16, 0
-    sat[2] = 32, 16, 0
+    delays = np.zeros((1, 5, 2, 2))
+    delays[0, 0, 0, 0] = 0.1  # A rise -> Z rise
+    delays[0, 0, 0, 1] = 0.2  # A rise -> Z fall
+    delays[0, 0, 1, 0] = 0.1  # A fall -> Z rise
+    delays[0, 0, 1, 1] = 0.2  # A fall -> Z fall
+    delays[0, 1, :, 0] = 0.3  # as above for B -> Z
+    delays[0, 1, :, 1] = 0.4
+    delays[0, 2, :, 0] = 0.5  # as above for C -> Z
+    delays[0, 2, :, 1] = 0.6
+    delays[0, 3, :, 0] = 0.7  # as above for D -> Z
+    delays[0, 3, :, 1] = 0.8
 
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert TMIN == z[0]
+    simctl_int = np.asarray([0], dtype=np.int32)
+    simctl_float = np.asarray([1], dtype=np.float32)
 
-    a[0] = TMIN
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert TMIN == z[0]
-    
-    b[0] = TMIN
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert TMAX == z[0]
+    def wave_assert(inputs, output):
+        for i, a in zip(inputs, c.reshape(-1,16)): a[:len(i)] = i
+        wave_eval_cpu(op, c, c_locs, c_caps, ebuf, 0, delays, simctl_int, simctl_float, 0, 0)
+        for i, v in enumerate(output): np.testing.assert_allclose(c.reshape(-1,16)[4,i], v)
 
-    a[0] = 1  # A _/^^^
-    b[0] = 2  # B __/^^
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert TMIN == z[0]  # ^^^\___ B -> Z fall delay
-    assert 2.4 == z[1]
-    assert TMAX == z[2]
+    wave_assert([[TMAX,TMAX],[TMAX,TMAX],[TMIN,TMAX],[TMIN,TMAX]], [TMIN,TMAX]) # NAND(0,0,1,1) => 1
+    wave_assert([[TMIN,TMAX],[TMAX,TMAX],[TMIN,TMAX],[TMIN,TMAX]], [TMIN,TMAX]) # NAND(1,0,1,1) => 1
+    wave_assert([[TMIN,TMAX],[TMIN,TMAX],[TMIN,TMAX],[TMIN,TMAX]], [TMAX])      # NAND(1,1,1,1) => 0
 
-    a[0] = TMIN  # A ^^^^^^
-    b[0] = TMIN  # B ^^^\__
-    b[1] = 2
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert 2.3 == z[0]  # ___/^^^ B -> Z rise delay
-    assert TMAX == z[1]
-
-    # pos pulse of 0.35 at B -> 0.45 after delays
-    a[0] = TMIN  # A ^^^^^^^^
-    b[0] = TMIN
-    b[1] = 2     # B ^^\__/^^
-    b[2] = 2.35
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert 2.3 == z[0]  # __/^^\__
-    assert 2.75 == z[1]
-    assert TMAX == z[2]
-
-    # neg pulse of 0.45 at B -> 0.35 after delays
-    a[0] = TMIN  # A ^^^^^^^^
-    b[0] = 2  # B __/^^\__
-    b[1] = 2.45
-    b[2] = TMAX
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert TMIN == z[0]  # ^^\__/^^
-    assert 2.4 == z[1]
-    assert 2.75 == z[2]
-    assert TMAX == z[3]
-
-    # neg pulse of 0.35 at B -> 0.25 after delays (filtered)
-    a[0] = TMIN  # A ^^^^^^^^
-    b[0] = 2  # B __/^^\__
-    b[1] = 2.35
-    b[2] = TMAX
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert TMIN == z[0]  # ^^^^^^
-    assert TMAX == z[1]
-
-    # pos pulse of 0.25 at B -> 0.35 after delays (filtered)
-    a[0] = TMIN  # A ^^^^^^^^
-    b[0] = TMIN
-    b[1] = 2  # B ^^\__/^^
-    b[2] = 2.25
-    wave_eval((0b0111, 2, 0, 1), state, sat, 0, line_times)
-    assert TMAX == z[0]  # ______
+    # Keep inputs C=1 and D=1.
+    wave_assert([[1,TMAX],[2,TMAX]], [TMIN,2.4,TMAX])              # _/⎺⎺⎺ NAND __/⎺⎺ => ⎺⎺⎺\___ (B->Z fall delay)
+    wave_assert([[TMIN,TMAX],[TMIN,2,TMAX]],  [2.3,TMAX])          # ⎺⎺⎺⎺⎺ NAND ⎺⎺\__ => ___/⎺⎺⎺ (B->Z rise delay)
+    wave_assert([[TMIN,TMAX],[TMIN,2,2.35,TMAX]], [2.3,2.75,TMAX]) # ⎺⎺⎺⎺⎺ NAND ⎺\_/⎺ => __/⎺⎺\_ (pos pulse, .35@B -> .45@Z)
+    wave_assert([[TMIN,TMAX],[TMIN,2,2.25,TMAX]], [TMAX])          # ⎺⎺⎺⎺⎺ NAND ⎺\_/⎺ => _______ (pos pulse, .25@B -> .35@Z, filtered)
+    wave_assert([[TMIN,TMAX],[2,2.45,TMAX]], [TMIN,2.4,2.75,TMAX]) # ⎺⎺⎺⎺⎺ NAND _/⎺\_ => ⎺⎺\_/⎺⎺ (neg pulse, .45@B -> .35@Z)
+    wave_assert([[TMIN,TMAX],[2,2.35,TMAX]], [TMIN,TMAX])          # ⎺⎺⎺⎺⎺ NAND _/⎺\_ => ⎺⎺⎺⎺⎺⎺⎺ (neg pulse, .35@B -> .25@Z, filtered)
 
 
-def compare_to_logic_sim(wsim):
-    tests = PackedVectors(wsim.sims, len(wsim.interface), 3)
-    tests.randomize()
-    wsim.assign(tests)
-    wsim.propagate(8)
-    cdata = wsim.capture()
+def test_tiny_circuit():
+    c = bench.parse('input(x, y) output(a, o, n) a=and(x,y) o=or(x,y) n=not(x)')
+    delays = np.full((1, len(c.lines), 2, 2), 1.0)  # unit delay for all lines
+    wsim = WaveSim(c, delays)
+    assert wsim.s.shape[1] == 5
 
-    resp = tests.copy()
+    # values for x
+    wsim.s[:3,0,0] = 0, 10, 0
+    wsim.s[:3,0,1] = 0, 20, 1
+    wsim.s[:3,0,2] = 1, 30, 0
+    wsim.s[:3,0,3] = 1, 40, 1
 
-    for iidx, inode in enumerate(wsim.interface):
-        if len(inode.ins) > 0:
-            for vidx in range(wsim.sims):
-                resp.set_value(vidx, iidx, 0 if cdata[iidx, vidx, 0] < 0.5 else 1)
+    # values for y
+    wsim.s[:3,1,0] = 1, 50, 0
+    wsim.s[:3,1,1] = 1, 60, 0
+    wsim.s[:3,1,2] = 1, 70, 0
+    wsim.s[:3,1,3] = 0, 80, 1
 
-    lsim = LogicSim(wsim.circuit, len(tests), 3)
-    lsim.assign(tests)
-    lsim.propagate()
-    exp = tests.copy()
-    lsim.capture(exp)
-    
-    for i in range(8):
-        exp_str = exp[i].replace('R', '1').replace('F', '0').replace('P', '0').replace('N', '1')
-        res_str = resp[i].replace('R', '1').replace('F', '0').replace('P', '0').replace('N', '1')
-        assert res_str == exp_str
+    wsim.s_to_c()
+
+    x_c_loc = wsim.c_locs[wsim.ppi_offset+0] # check x waveforms
+    np.testing.assert_allclose(wsim.c[x_c_loc:x_c_loc+3, 0], [TMAX, TMAX, TMAX])
+    np.testing.assert_allclose(wsim.c[x_c_loc:x_c_loc+3, 1], [20, TMAX, TMAX])
+    np.testing.assert_allclose(wsim.c[x_c_loc:x_c_loc+3, 2], [TMIN, 30, TMAX])
+    np.testing.assert_allclose(wsim.c[x_c_loc:x_c_loc+3, 3], [TMIN, TMAX, TMAX])
+
+    y_c_loc = wsim.c_locs[wsim.ppi_offset+1] # check y waveforms
+    np.testing.assert_allclose(wsim.c[y_c_loc:y_c_loc+3, 0], [TMIN, 50, TMAX])
+    np.testing.assert_allclose(wsim.c[y_c_loc:y_c_loc+3, 1], [TMIN, 60, TMAX])
+    np.testing.assert_allclose(wsim.c[y_c_loc:y_c_loc+3, 2], [TMIN, 70, TMAX])
+    np.testing.assert_allclose(wsim.c[y_c_loc:y_c_loc+3, 3], [80, TMAX, TMAX])
+
+    wsim.c_prop()
+
+    a_c_loc = wsim.c_locs[wsim.ppo_offset+2] # check a waveforms
+    np.testing.assert_allclose(wsim.c[a_c_loc:a_c_loc+3, 0], [TMAX, TMAX, TMAX])
+    np.testing.assert_allclose(wsim.c[a_c_loc:a_c_loc+3, 1], [21, 61, TMAX])
+    np.testing.assert_allclose(wsim.c[a_c_loc:a_c_loc+3, 2], [TMIN, 31, TMAX])
+    np.testing.assert_allclose(wsim.c[a_c_loc:a_c_loc+3, 3], [81, TMAX, TMAX])
+
+    o_c_loc = wsim.c_locs[wsim.ppo_offset+3] # check o waveforms
+    np.testing.assert_allclose(wsim.c[o_c_loc:o_c_loc+3, 0], [TMIN, 51, TMAX])
+    np.testing.assert_allclose(wsim.c[o_c_loc:o_c_loc+3, 1], [TMIN, TMAX, TMAX])
+    np.testing.assert_allclose(wsim.c[o_c_loc:o_c_loc+3, 2], [TMIN, 71, TMAX])
+    np.testing.assert_allclose(wsim.c[o_c_loc:o_c_loc+3, 3], [TMIN, TMAX, TMAX])
+
+    n_c_loc = wsim.c_locs[wsim.ppo_offset+4] # check n waveforms
+    np.testing.assert_allclose(wsim.c[n_c_loc:n_c_loc+3, 0], [TMIN, TMAX, TMAX])
+    np.testing.assert_allclose(wsim.c[n_c_loc:n_c_loc+3, 1], [TMIN, 21, TMAX])
+    np.testing.assert_allclose(wsim.c[n_c_loc:n_c_loc+3, 2], [31, TMAX, TMAX])
+    np.testing.assert_allclose(wsim.c[n_c_loc:n_c_loc+3, 3], [TMAX, TMAX, TMAX])
+
+    wsim.c_to_s()
+
+    # check a captures
+    np.testing.assert_allclose(wsim.s[3:7, 2, 0], [0, TMAX, TMIN, 0])
+    np.testing.assert_allclose(wsim.s[3:7, 2, 1], [0, 21, 61, 0])
+    np.testing.assert_allclose(wsim.s[3:7, 2, 2], [1, 31, 31, 0])
+    np.testing.assert_allclose(wsim.s[3:7, 2, 3], [0, 81, 81, 1])
+
+    # check o captures
+    np.testing.assert_allclose(wsim.s[3:7, 3, 0], [1, 51, 51, 0])
+    np.testing.assert_allclose(wsim.s[3:7, 3, 1], [1, TMAX, TMIN, 1])
+    np.testing.assert_allclose(wsim.s[3:7, 3, 2], [1, 71, 71, 0])
+    np.testing.assert_allclose(wsim.s[3:7, 3, 3], [1, TMAX, TMIN, 1])
+
+    # check o captures
+    np.testing.assert_allclose(wsim.s[3:7, 4, 0], [1, TMAX, TMIN, 1])
+    np.testing.assert_allclose(wsim.s[3:7, 4, 1], [1, 21, 21, 0])
+    np.testing.assert_allclose(wsim.s[3:7, 4, 2], [0, 31, 31, 1])
+    np.testing.assert_allclose(wsim.s[3:7, 4, 3], [0, TMAX, TMIN, 0])
 
 
-def test_b14(mydir):
-    c = verilog.parse(mydir / 'b14.v.gz', branchforks=True)
-    df = sdf.parse(mydir / 'b14.sdf.gz')
-    lt = df.annotation(c, pin_index)
-    wsim = WaveSim(c, lt, 8)
-    compare_to_logic_sim(wsim)
+def compare_to_logic_sim(wsim: WaveSim):
+    choices = np.asarray([logic.ZERO, logic.ONE, logic.RISE, logic.FALL], dtype=np.uint8)
+    rng = np.random.default_rng(10)
+    tests = rng.choice(choices, (wsim.s_len, wsim.sims))
+
+    wsim.s[0] = (tests & 2) >> 1
+    wsim.s[3] = (tests & 2) >> 1
+    wsim.s[1] = 0.0
+    wsim.s[2] = tests & 1
+    wsim.s[6] = tests & 1
+
+    wsim.s_to_c()
+    wsim.c_prop()
+    wsim.c_to_s()
+
+    resp = np.array(wsim.s[6], dtype=np.uint8) | (np.array(wsim.s[3], dtype=np.uint8)<<1)
+    resp |= ((resp ^ (resp >> 1)) & 1) << 2  # transitions
+    resp[wsim.pi_s_locs] = logic.UNASSIGNED
+
+    lsim = LogicSim6V(wsim.circuit, tests.shape[-1])
+    lsim.s_assign[...] = tests
+    lsim.s_to_c()
+    lsim.c_prop()
+    lsim.c_to_s()
+    exp = lsim.s_result[:,:tests.shape[-1]]
+
+    resp[resp == logic.PPULSE] = logic.ZERO
+    resp[resp == logic.NPULSE] = logic.ONE
+
+    exp[exp == logic.PPULSE] = logic.ZERO
+    exp[exp == logic.NPULSE] = logic.ONE
+
+    np.testing.assert_allclose(resp, exp)
 
 
-def test_b14_strip_forks(mydir):
-    c = verilog.parse(mydir / 'b14.v.gz', branchforks=True)
-    df = sdf.parse(mydir / 'b14.sdf.gz')
-    lt = df.annotation(c, pin_index)
-    wsim = WaveSim(c, lt, 8, strip_forks=True)
-    compare_to_logic_sim(wsim)
+def test_b15(b15_2ig_circuit_resolved, b15_2ig_delays):
+    compare_to_logic_sim(WaveSim(b15_2ig_circuit_resolved, b15_2ig_delays, 8))
 
 
-def test_b14_cuda(mydir):
-    c = verilog.parse(mydir / 'b14.v.gz', branchforks=True)
-    df = sdf.parse(mydir / 'b14.sdf.gz')
-    lt = df.annotation(c, pin_index)
-    wsim = WaveSimCuda(c, lt, 8)
-    compare_to_logic_sim(wsim)
+def test_b15_strip_forks(b15_2ig_circuit_resolved, b15_2ig_delays):
+    compare_to_logic_sim(WaveSim(b15_2ig_circuit_resolved, b15_2ig_delays, 8, strip_forks=True))
+
+
+def test_b15_cuda(b15_2ig_circuit_resolved, b15_2ig_delays):
+    compare_to_logic_sim(WaveSimCuda(b15_2ig_circuit_resolved, b15_2ig_delays, 8, strip_forks=True))