       j++;
     }
     i++;
+  }
+
+  __syncthreads();
+  if (lid == 0) {
+    final[0] = A[0][0]; final[1] = A[0][1];
+    final[2] = A[0][2]; final[3] = A[0][3];
+    final[4] = A[1][0]; final[5] = A[1][1];
+    final[6] = A[1][2]; final[7] = A[1][3];
   }
 }
 

} while(0);

//set x and y through xyvals
__global__ void k2(int *xyvals, int *trace) {
  __shared__ int A[2][4];
  int buf, x, y, i, j;
  int lid = threadIdx.x;

    printf("---\n");
  }

  // print out final state
  cudaMemcpy(final, d_final, d_final_size, cudaMemcpyDeviceToHost);
  printf("final state\n");
  printf("    A = {{%d,%d,%d,%d}, {%d,%d,%d,%d}}\n",
    final[0],final[1],final[2],final[3],
    final[4],final[5],final[6],final[7]);

  cudaFree(d_xyvals);
  cudaFree(d_trace);
  cudaFree(d_final);
  delete[] trace;
  return 0;

