__kernel void ecore(__global int column, __global int row)
{
  ecore_t* core = &cores[column][row];

  while (1)
  {
    if(ACTIVE)
    {
      int32_t op = read_memory32(core, PC);

      if(PC & 0x01) {
        uint32_t status = read_memory32(core, MEM_REG_STATUS);
        status |= 0x00070000;                   // set EXCAUSE, don't know the exact value so I used 111b

        PC+=1;                                  // align to next word boundary

        raise_interrupt(INT_SOFTWARE_EXCEPTION);// basically reset the core
      }

      if(op & 0x01 || op & 0xF == 0xE || op & 0xF == 0xA)
        op_16bit(op & 0x0000FFFF, core);
      else
        op_32bit(op, core);

    }
  }
}