 
 void *kernelInit() {
 	KernelContext *ctx = (KernelContext*) calloc(1, sizeof(KernelContext));
+	cudaMalloc(&ctx->dbacklogs, sizeof(ctx->backlogs));
+	cudaMalloc(&ctx->ddiffs, sizeof(ctx->diffs));
 	return ctx;
 }
 

	for (int i = 0; i < NDIFFS; i++) {
		cudaFree(ctx->diffs[i]);
	}
	free(ctx);
}

}

__global__ static void updateDiff(
		uint32_t *d, uint32_t *b0, uint32_t *b1, int w) {
	if (!(b0 && b1)) {
		return;
	}
	int k = (blockDim.x * blockIdx.x + threadIdx.x) * w +
		blockDim.y * blockIdx.y + threadIdx.y;
	d[k] = diff(b0[k], b1[k]);
}

#define in(i, j) in[(i)*w + (j)]
