'ptxas' died with status 0xC0000005 (ACCESS_VIOLAT

Hello, guys,

I’ve been looking all over the internet to find the solution to this problem. I am getting ‘ptxas’ died with status 0xC0000005 (ACCESS_VIOLATION) linker error when I am trying to compile my program. I am doing SHA 256 implementation and here is the excerpt:

#define PUT_UINT32_BE(n,b,i)      \

{            	\

    (b)[(i)    ] = (unsigned char) ( (n) >> 24 );	\

    (b)[(i) + 1] = (unsigned char) ( (n) >> 16 );	\

    (b)[(i) + 2] = (unsigned char) ( (n) >>  8 );	\

    (b)[(i) + 3] = (unsigned char) ( (n)       );	\


typedef struct {

	SHA256_LONG state[8];	/* Intermediate hash state */

} sha256_context;

__device__ void sha256_finish (sha256_context *ctx, unsigned char *output)


	int i;

	for (i = 0; i <8; i++)

  PUT_UINT32_BE(ctx->state[i], output, i * 4);


__device__ void sha256_run (SHA256_CHAR *input, unsigned char *output)


	sha256_context ctx;

	sha256_start (&ctx);

	sha256_compute (&ctx, input);

	sha256_finish (&ctx, output);


__global__ void sha256_kernel (unsigned char *input, unsigned int size, unsigned char *output, int totalThreads)


	int threadIndex = threadIdx.x + blockDim.x * blockIdx.x;

	if (threadIndex > totalThreads - 1)


	/* Is this the last thread? If so, we have to compute hash for current block and padding, if any. */

	if (threadIndex == totalThreads - 1) {

  /* patikrinti ar nera viskas suhashinta, nes gali buti taip kad */

 /* we check if message size is multiple of block size */

  if (size % 64 == 0) {

  	/* here we need to hash only padding */

  	/* cia yra kaikas blogai, nes message len yra tik 32 bitu */

  	/*memcpy(sha256_padding + 60, &size, 8);

  	sha256_run (sha256_padding, output + threadIndex * SHA256_HASH_SIZE);*/

  } else {

  	/* here we have to calculate how big padding has to be

    * in worst case we might need to hash two more blocks */




	sha256_run (input + threadIndex * SHA256_BLOCK_SIZE, output + threadIndex * SHA256_HASH_SIZE);


I just don’t know what the problem is. If I comment out the loop in sha256_finish function I won’t get any compilation errors. Can anyone help me? Thank you.


I ran into a compiler bug myself a while back. Tracking it down was fairly straight-forward. The routine was fairly small, so I just went through and commented the first line and tried to compile. Then uncomment the first line and comment the second, and compile, and so on, until the problem went away. In my particular case, I was defining a variable and assigning it on the same line which normally isn’t a problem. I broke the line into 2 statements: definition and assignment, and then it compiled fine.

Once you track down the problem, there are various things you can try. Re-ordering things near that code, if possible, might help. Or, add a line of code before the line that’s failing. Something like a simple assignment that doesn’t do anything.

Sometimes compilers build up an invalid state from previous code and they fail. Generally, re-ordering the code or adding or removing code before the problem area, can change the state so that it’s no longer invalid.

I ran into a similar bug in the 16-bit Visual C++ compiler back in the day…

This looks like a compiler bug. If you can attach a test app which reproduces the problem, we can investigate futher.

Oh, I am sorry guys, that was such a lame mistake I was doing all the time. There was something like this: memcpy(buffer, &size, 8) where size is unsigned integer and I am trying to copy 8 bytes. I am still attaching the code if you are interested.

CUDAProject.zip (34.1 KB)