PTXAS died with status 0xC0000005 (ACCES VIOLATION Compiler problem?

Accelerated Computing CUDA CUDA Programming and Performance
1

Initially I thought it was my code issue. A search through the forum revealed that there r others who have come across this.

I have got a very simple kernel that reproduces this problem. I am in a CUDA 1.1 environment and I am NOT the “administrator” of my system.

#include <cuda.h>

#include <stdio.h>

struct texture <float> texrefFloat;

void *devptr;

__global__ void mykernel(texture <float> ref, float *g_out)

{

	float data;

	data = tex1Dfetch(ref, threadIdx.x);

	g_out[threadIdx.x] = data+1;

	return;

}

int output[32];

int main()

{

	dim3 grid, block;

	devptr = NULL;

	cudaMalloc(&devptr, 100*sizeof(float));

	if (devptr == NULL)

	{

  printf("cudaMalloc() failed.\n");

  return -1;

	}

	cudaBindTexture(0, texrefFloat, devptr, 100*sizeof(float));

	cudaMemset(devptr, 0, 100*sizeof(float));

	block.x = 32;

	// mykernel <<< grid, block >>> (texrefFloat, (float *)devptr);

	//cudaThreadSynchronize();

	// cudaMemcpy(output, devptr, 32*sizeof(float), cudaMemcpyDeviceToHost);

	return 0;

}

I am just learning textures. So, Please excuse me if its the mistake of my program.

When I compile this code, I get the message:

"

1>------ Build started: Project: texture, Configuration: Debug Win32 ------

1>Performing Custom Build Step

1>tex.cu

1>tex.cudafe1.gpu

1>tex.cudafe2.gpu

1>Internal error

1>nvcc error : ‘ptxas’ died with status 0xC0000005 (ACCESS_VIOLATION)

1>Linking…

1>LINK : fatal error LNK1181: cannot open input file ‘.\Debug\tex.obj’

1>Build log was saved at “file://c:\CUDA\sarnath\texture\Debug\BuildLog.htm”

1>texture - 1 error(s), 0 warning(s)

========== Build: 0 succeeded, 1 failed, 0 up-to-date, 0 skipped ==========

"

2

You can’t use textures as parameters of your kernel, just remove the parameter and use texrefFloat and you should be fine.

3

Note that if I comment out the “g_out[threadIdx.x] = data+1;” line from my code, the compilation goes fine.

Also “g_out[threadIdx.x] = 5;” works!

4

aah. Thanks. That worked!! I thought that could be it. But since it did NOT have the “device” qualifier, I thought I could NOT use it inside my kernel… Mmm… The CUDA grammar is context sensitive… Durrrr…