[driver api] Can not get the right values form 3d texture... texture3d, driver api

Hi, I’m learning CUDA from GPU Computing SDK 6.0 and trying to write simpleTexture3D with driver api.(ubnuntu 12.04-32, GTX 660)

Right now I can create a 3d CUarray and bind a CUtexref on it, all with the CUDA_SUCCESS result. But the values from tex3D function in cu file are always zero.

but 1d texture is the correct value…What’s the problem?

Here are the codes, please help, thanks!

float *input_float_3D = (float *)malloc(sizeof(float)*2*2*2);
   for (int x = 0; x < 2; x++){
            for (int y = 0; y < 2; y++){
                int xy = x + y * 2;
                for (int z = 0; z < 2; z++){
                    int xyz = xy + z * 2 * 2;
                    input_float_3D[xyz] = xyz;
                }
            }
        }

	for(int i=0; i<2*2*2; i++){
	   printf("%d  %f\n",i, input_float_3D[i]);
   }
   
   CUarray array2;
   CUDA_ARRAY3D_DESCRIPTOR ad2;
   ad2.Format = CU_AD_FORMAT_FLOAT;
   ad2.Width = 2;
   ad2.Height = 2;
   ad2.Depth = 2;
   ad2.Flags=1;
   ad2.NumChannels = 1;
   CUresult error3 =cuArray3DCreate(&array2, &ad2);
   
   CUDA_MEMCPY3D copy;
   memset(&copy, 0, sizeof(copy));
   copy.srcMemoryType = CU_MEMORYTYPE_HOST;
   copy.srcHost = input_float_3D;
   copy.srcPitch = 2;
   copy.Height = 2;
   
   copy.dstMemoryType = CU_MEMORYTYPE_ARRAY;
   copy.dstArray = array2;
   
   copy.WidthInBytes = 2;
   copy.Height = 2;
   copy.Depth = 2;
  
   CUresult error4 =  cuMemcpy3D(&copy);
   
   CUtexref texref2;
   CUresult error5 = cuModuleGetTexRef(&texref2, pmodule->m_module, "texture_float_3D");
   CUresult error6 = cuTexRefSetFilterMode(texref2, CU_TR_FILTER_MODE_POINT);
   CUresult error7 =cuTexRefSetAddressMode(texref2, 0, CU_TR_ADDRESS_MODE_CLAMP);
   CUresult error8 =cuTexRefSetAddressMode(texref2, 1, CU_TR_ADDRESS_MODE_CLAMP);
   CUresult error9 =cuTexRefSetAddressMode(texref2, 2, CU_TR_ADDRESS_MODE_CLAMP);
   CUresult error10 =cuTexRefSetFlags(texref2, CU_TRSF_NORMALIZED_COORDINATES);
   CUresult error11 =cuTexRefSetFormat(texref2, CU_AD_FORMAT_FLOAT, 1);
   CUresult error12 =cuTexRefSetArray(texref2, array2, CU_TRSA_OVERRIDE_FORMAT);

result->Set(String::New("cuArray3DCreate error"), Integer::New(error3));
   result->Set(String::New("cuMemcpy3D error"), Integer::New(error4));
    result->Set(String::New("cuModuleGetTexRef error"), Integer::New(error5));
    result->Set(String::New("cuTexRefSetFilterMode error"), Integer::New(error6));
    result->Set(String::New("cuTexRefSetAddressMode1 error"), Integer::New(error7));
    result->Set(String::New("cuTexRefSetAddressMode2 error"), Integer::New(error8));
    result->Set(String::New("cuTexRefSetAddressMode3 error"), Integer::New(error9));
    result->Set(String::New("cuTexRefSetFlags error"), Integer::New(error10));
    result->Set(String::New("cuTexRefSetFormat error"), Integer::New(error11));
    result->Set(String::New("cuTexRefSetArray error"), Integer::New(error12));
texture<float,  3, cudaReadModeElementType> texture_float_3D;
extern "C" {
__global__ void helloWorld( float *data3) {

        float result = tex3D(texture_float_3D, 0.0f,0.0f,0.0f);
       data3[0] = result;

please help me