C# CudaDeviceVariable类代码示例

 /// <summary>
 /// Forward DCT, quantization and level shift part of the JPEG encoding.
 /// Input is expected in 8x8 macro blocks and output is expected to be in 64x1
 /// macro blocks.
 /// </summary>
 /// <param name="src">Source image.</param>
 /// <param name="dst">Destination image</param>
 /// <param name="QuantFwdTable">Forward quantization tables for JPEG encoding created using QuantInvTableInit()</param>
 /// <param name="oSizeRoi">Roi size (in macro blocks?).</param>
 public static void DCTQuantFwd8x8LS(NPPImage_8uC1 src, NPPImage_16sC1 dst, CudaDeviceVariable<ushort> QuantFwdTable, NppiSize oSizeRoi)
 {
     NppStatus status;
     status = NPPNativeMethods.NPPi.ImageCompression.nppiDCTQuantFwd8x8LS_JPEG_8u16s_C1R(src.DevicePointer, src.Pitch, dst.DevicePointer, dst.Pitch, QuantFwdTable.DevicePointer, oSizeRoi);
     Debug.WriteLine(String.Format("{0:G}, {1}: {2}", DateTime.Now, "nppiDCTQuantFwd8x8LS_JPEG_8u16s_C1R", status));
     NPPException.CheckNppStatus(status, null);
 }

        protected override void Init()
        {
            var kernelFileName = KernelFile;
            var initKernel = Ctx.LoadKernel(kernelFileName, "generateData");
            B = new CudaDeviceVariable<double>(DimensionsCount);
            M = new CudaDeviceVariable<double>(DimensionsCount * DimensionsCount);
            Xopt = new CudaDeviceVariable<double>(DimensionsCount);

            var d_fopt = new CudaDeviceVariable<double>(1);

            long rseed = FunctionNumber + 10000 * InstanceNumber;

            initKernel.Run(
                DimensionsCount,
                rseed,
                FunctionNumber,
                InstanceNumber,
                M.DevicePointer,
                B.DevicePointer,
                Xopt.DevicePointer,
                d_fopt.DevicePointer);

            double[] fopt_arr = d_fopt;

            Fopt = fopt_arr[0];
        }

        public CudaIntersectionDevice(RayEngineScene scene, NVContext ctx)
            : base(scene)
        {
            wallclock = new Stopwatch();
            this.todoRayBuffers = new ConcurrentQueue<Tuple<int, RayBuffer>>();
            this.doneRayBuffers = new List<ConcurrentQueue<RayBuffer>>() { { new ConcurrentQueue<RayBuffer>() } };
            this.started = false;
            if (ctx != null)
            {
                this.cudaContext = ctx;
            }
            else
            {
                this.cudaContext = new NVContext() { Context = new CudaContext(CudaContext.GetMaxGflopsDeviceId()) };
            }
            using (var sr = new StreamReader(@"G:\Git\RayDen\CudaMegaRay\x64\Release\kernel.cu.ptx"))
            {
                intersectKernel = cudaContext.Context.LoadKernelPTX(sr.BaseStream, "IntersectLBvh");
            }

            this.rays = new CudaDeviceVariable<RayData>(RayBuffer.RayBufferSize);
            this.hits = new CudaDeviceVariable<RayHit>(RayBuffer.RayBufferSize);
            verts = scene.Vertices.ToArray();
            //scene.Triangles.Select(i => i.GetInfo()).ToArray();

            var ti = scene.Triangles.Select(i => i.GetInfo()).ToArray();
            var da = new BvhDataAdapter(scene);
            var treeData = da.GetMpData();
            bvh = treeData;
            trianglesCount = ti.Length;
            tris = ti; 

            nodesCount = treeData.Length;
            Tracer.TraceLine("BVH Data Size {0:F3} MBytes", (treeData.Length * 32f) / (1024f * 1024f));
        }

        protected override void Init()
        {
            var kernelFileName = KernelFile;
            var initKernel = Ctx.LoadKernel(kernelFileName, "generateData");
            Xopt = new CudaDeviceVariable<double>(DimensionsCount);

            var d_fopt = new CudaDeviceVariable<double>(1);

            long rseed = FunctionNumber + 10000 * InstanceNumber;

            double[] host_minus_ones = new double[DimensionsCount];

            for (int i = 0; i < DimensionsCount; i++)
            {
                host_minus_ones[i] = -1;
            }

            MinusOnes = host_minus_ones;

            Factor = Math.Max(1.0, Math.Sqrt(DimensionsCount)/8.0);

            initKernel.Run(
                DimensionsCount,
                rseed,
                FunctionNumber,
                InstanceNumber,
                Xopt.DevicePointer,
                d_fopt.DevicePointer);

            double[] fopt_arr = d_fopt;

            Fopt = fopt_arr[0];
        }

        public void Backward(CudnnPoolingDescriptor pooling, CudnnTensorDescriptor srcTensor, float[] srcData, CudnnTensorDescriptor srcDiffTensor, float[] srcDiffData,
                                                             CudnnTensorDescriptor destTensor, float[] destData, CudnnTensorDescriptor destDiffTensor, float[] destDiffData)
        {
            Contract.Requires(pooling != null);
            Contract.Requires(srcTensor != null);
            Contract.Requires(srcData != null);
            Contract.Requires(destTensor != null);
            Contract.Requires(destData != null);
            Contract.Requires(srcDiffTensor != null);
            Contract.Requires(srcDiffData != null);
            Contract.Requires(destDiffTensor != null);
            Contract.Requires(destDiffData != null);

            ThrowIfNotInitialized();
            CheckIfCompatible(CudnnType.Float, srcTensor, srcDiffTensor, destTensor, destDiffTensor);

            using (var srcDataGpu = new CudaDeviceVariable<float>(srcData.Length))
            using (var srcDiffDataGpu = new CudaDeviceVariable<float>(srcDiffData.Length))
            using (var destDataGpu = new CudaDeviceVariable<float>(destData.Length))
            using (var destDiffDataGpu = new CudaDeviceVariable<float>(destDiffData.Length))
            {
                srcDataGpu.CopyToDevice(srcData);
                srcDiffDataGpu.CopyToDevice(srcDiffData);
                destDataGpu.CopyToDevice(destData);

                Invoke(() => CudnnNativeMethods.cudnnPoolingBackward(handle, pooling.Handle,
                                                                     srcTensor.Handle, srcDataGpu.DevicePointer, srcDiffTensor.Handle, srcDiffDataGpu.DevicePointer,
                                                                     destTensor.Handle, destDataGpu.DevicePointer, destDiffTensor.Handle, destDiffDataGpu.DevicePointer));
                destDiffDataGpu.CopyToHost(destDiffData);
            }
        }

        ////////////////////////////////////////////////////////////////////////////////
        // Occupancy-based launch configurator
        //
        // The launch configurator, cudaOccupancyMaxPotentialBlockSize and
        // cudaOccupancyMaxPotentialBlockSizeVariableSMem, suggests a block
        // size that achieves the best theoretical occupancy. It also returns
        // the minimum number of blocks needed to achieve the occupancy on the
        // whole device.
        //
        // This launch configurator is purely occupancy-based. It doesn't
        // translate directly to performance, but the suggestion should
        // nevertheless be a good starting point for further optimizations.
        //
        // This function configures the launch based on the "automatic"
        // argument, records the runtime, and reports occupancy and runtime.
        ////////////////////////////////////////////////////////////////////////////////
        static int launchConfig(CudaDeviceVariable<int> array, int arrayCount, bool automatic)
        {
            int blockSize = 0;
            int minGridSize = 0;
            int gridSize;
            SizeT dynamicSMemUsage = 0;

            float elapsedTime;

            double potentialOccupancy;

            CudaOccupancy.cudaOccDeviceState state = new CudaOccupancy.cudaOccDeviceState();
            state.cacheConfig = CudaOccupancy.cudaOccCacheConfig.PreferNone;

            if (automatic)
            {
                CudaOccupancy.cudaOccMaxPotentialOccupancyBlockSize(ref minGridSize, ref blockSize, new CudaOccupancy.cudaOccDeviceProp(0), new CudaOccupancy.cudaOccFuncAttributes(kernel), state, dynamicSMemUsage);

                Console.WriteLine("Suggested block size: {0}", blockSize);
                Console.WriteLine("Minimum grid size for maximum occupancy: {0}", minGridSize);
            }
            else
            {
                // This block size is too small. Given limited number of
                // active blocks per multiprocessor, the number of active
                // threads will be limited, and thus unable to achieve maximum
                // occupancy.
                //
                blockSize = manualBlockSize;
            }

            // Round up
            //
            gridSize = (arrayCount + blockSize - 1) / blockSize;

            // Launch and profile
            //
            kernel.GridDimensions = gridSize;
            kernel.BlockDimensions = blockSize;
            elapsedTime = kernel.Run(array.DevicePointer, arrayCount);

            // Calculate occupancy
            //
            potentialOccupancy = reportPotentialOccupancy(blockSize, dynamicSMemUsage);

            Console.WriteLine("Potential occupancy: {0}%", potentialOccupancy * 100);

            // Report elapsed time
            //
            Console.WriteLine("Elapsed time: {0}ms", elapsedTime * 100);

            return 0;
        }

        public void BackwardBias(CudnnTensorDescriptor srcTensor, CudaDeviceVariable<double> srcData, CudnnTensorDescriptor destTensor, CudaDeviceVariable<double> destData, CudnnAccumulateResult accumulate)
        {
            Contract.Requires(srcTensor != null);
            Contract.Requires(srcData != null);
            Contract.Requires(destTensor != null);
            Contract.Requires(destData != null);

            ThrowIfNotInitialized();
            CheckIfCompatible(CudnnType.Double, srcTensor, destTensor);

            Invoke(() => CudnnNativeMethods.cudnnConvolutionBackwardBias(handle, srcTensor.Handle, srcData.DevicePointer, destTensor.Handle, destData.DevicePointer, accumulate));
        }

        public void BackwardData(CudnnFilterDescriptor filter, CudaDeviceVariable<float> filterData, CudnnTensorDescriptor diffTensor, CudaDeviceVariable<float> diffData, CudnnConvolutionDescriptor convolution, CudnnTensorDescriptor gradient, CudaDeviceVariable<float> gradientData, CudnnAccumulateResult accumulate)
        {
            Contract.Requires(filter != null);
            Contract.Requires(filterData != null);
            Contract.Requires(diffTensor != null);
            Contract.Requires(diffData != null);
            Contract.Requires(convolution != null);
            Contract.Requires(gradient != null);
            Contract.Requires(gradientData != null);

            ThrowIfNotInitialized();
            CheckIfCompatible(CudnnType.Float, filter, diffTensor, gradient);

            Invoke(() => CudnnNativeMethods.cudnnConvolutionBackwardData(handle, filter.Handle, filterData.DevicePointer, diffTensor.Handle, diffData.DevicePointer, convolution.Handle, gradient.Handle, gradientData.DevicePointer, accumulate));
        }

		/// <summary>
		/// 
		/// </summary>
		/// <param name="size">Graph size</param>
		public GraphCut4(NppiSize size)
		{
			_size = size;
			int bufferSize = 0;
			status = NPPNativeMethods.NPPi.ImageLabeling.nppiGraphcutGetSize(_size, ref bufferSize);
			Debug.WriteLine(String.Format("{0:G}, {1}: {2}", DateTime.Now, "nppiGraphcutGetSize", status));
			NPPException.CheckNppStatus(status, this);

			_buffer = new CudaDeviceVariable<byte>(bufferSize);

			_state = new NppiGraphcutState();
			status = NPPNativeMethods.NPPi.ImageLabeling.nppiGraphcutInitAlloc(_size, ref _state, _buffer.DevicePointer);
			Debug.WriteLine(String.Format("{0:G}, {1}: {2}", DateTime.Now, "nppiGraphcutInitAlloc", status));
			NPPException.CheckNppStatus(status, this);
		}

        public void Run(MatOperation operation, CudaDeviceVariable<float> A, int ACount, int AColumnHint, CudaDeviceVariable<float> B, int BCount, int BColumnHint, CudaDeviceVariable<float> Result, int ResultCount, int ResultColumnHint, float beta = 1.0f)
        {
            Result.Memset(BitConverter.ToUInt32(BitConverter.GetBytes(0.0f), 0));

            switch (operation)
            {
                case MatOperation.Multiplication:  // vectors/matrices have to be always in the correct dimesions!
                    if (BCount > 1 && ACount > 1 && BColumnHint == 1 && ACount / AColumnHint > 1 && BCount / BColumnHint == AColumnHint) //. A*vecB
                    {
                        MyCublasFactory.Instance.Gemv(Operation.Transpose,  // transpose beacuase it does Ax row wise if x is a row vector :D
                            AColumnHint, ACount / AColumnHint, 1.0f,
                            A, AColumnHint,
                            B, 1,
                            beta, Result, 1);
                    }
                    else if (ACount > 1 && BCount > 1 && ACount / AColumnHint == 1 && BColumnHint > 1 && BCount / BColumnHint == AColumnHint)  // vecA*B
                    {
                        MyCublasFactory.Instance.Gemv(Operation.NonTranspose,  // transpose beacuase it does Ax row wise if x is a row vector :D
                            BColumnHint, BCount / BColumnHint, 1.0f,
                            B, BColumnHint,
                            A, 1,
                            beta, Result, 1);
                    }
                    else if (ACount / AColumnHint == 1 && BColumnHint == 1 && ACount > 1 && BCount > 1) //. trans(vecA) * vecB
                    {
                        Run(MatOperation.DotProd, A, ACount, AColumnHint, B, BCount, BColumnHint, Result, ResultCount, ResultColumnHint, beta);
                    }
                    else if (ACount != 1 || BCount != 1)// A*B   matrix multiplication
                    {
                        MyCublasFactory.Instance.Gemm(Operation.NonTranspose, Operation.NonTranspose,
                            ACount / AColumnHint, BColumnHint, AColumnHint, 1.0f,
                            A, ACount / AColumnHint,
                            B, BCount / BColumnHint,
                            beta, Result, ResultColumnHint);
                    }
                    break;
                case MatOperation.DotProd:
                    MyCublasFactory.Instance.Gemv(Operation.Transpose,  // transpose beacuase it does Ax row wise if x is a row vector :D
                       ACount, 1, 1.0f,
                       A, ACount,
                       B, 1,
                       beta, Result, 1);
                    break;
                default:
                    MyLog.Writer.WriteLine(MyLogLevel.ERROR, "Trying to run cublas for undefined MatOperation");
                    break;
            }
        }

        public CudaArray3D GenerateUniformArray(int width, int height, int depth)
        {
            int count = width * height * depth;

            CudaDeviceVariable<float> randomVariable = new CudaDeviceVariable<float>(count);
            CudaArray3D randomArray = new CudaArray3D(CUArrayFormat.Float, width, height, depth, CudaArray3DNumChannels.One, CUDAArray3DFlags.None);

            randomDevice.SetPseudoRandomGeneratorSeed((ulong)DateTime.Now.Ticks);
            randomDevice.GenerateUniform32(randomVariable.DevicePointer, count);

            randomArray.CopyFromDeviceToThis(randomVariable.DevicePointer, sizeof(float));

            randomVariable.Dispose();

            return randomArray;
        }

        /* Function to perform backward activation  */
        public void ActivationBackward(cudnnActivationMode mode,
										float alpha,
										TensorDescriptor srcDesc,
										CudaDeviceVariable<float> srcData,
										TensorDescriptor srcDiffDesc,
										CudaDeviceVariable<float> srcDiffData,
										TensorDescriptor destDesc,
										CudaDeviceVariable<float> destData,
										float beta,
										TensorDescriptor destDiffDesc,
										CudaDeviceVariable<float> destDiffData
										)
        {
            res = CudaDNNNativeMethods.cudnnActivationBackward(_handle, mode, ref alpha, srcDesc.Desc, srcData.DevicePointer, srcDiffDesc.Desc, srcDiffData.DevicePointer, destDesc.Desc, destData.DevicePointer, ref beta, destDiffDesc.Desc, destDiffData.DevicePointer);
            Debug.WriteLine(String.Format("{0:G}, {1}: {2}", DateTime.Now, "cudnnActivationForward", res));
            if (res != cudnnStatus.Success) throw new CudaDNNException(res);
        }

        protected override void Init()
        {
            var kernelFileName = KernelFile;
            var initKernel = Ctx.LoadKernel(kernelFileName, "generateData");
            Rotation2 = new CudaDeviceVariable<double>(DimensionsCount * DimensionsCount);
            Rotation1 = new CudaDeviceVariable<double>(DimensionsCount * DimensionsCount);
            Xopt = new CudaDeviceVariable<double>(DimensionsCount);

            long rseed = FunctionNumber + 10000 * InstanceNumber;

            initKernel.Run(
                DimensionsCount,
                rseed,
                Rotation1.DevicePointer,
                Rotation2.DevicePointer,
                Xopt.DevicePointer);
        }

        public static void DrawStringFromGPUMem(CudaDeviceVariable<float> inString, int x, int y, uint bgColor, uint fgColor, CUdeviceptr image, int imageWidth, int imageHeight, int stringOffset, int stringLen)
        {
            MyCudaKernel m_drawDigitKernel = MyKernelFactory.Instance.Kernel(MyKernelFactory.Instance.DevCount - 1, @"Observers\DrawStringKernel");
            CudaDeviceVariable<float> characters = MyMemoryManager.Instance.GetGlobalVariable<float>("CHARACTERS_TEXTURE", MyKernelFactory.Instance.DevCount - 1, LoadDigits);

            //MyKernelFactory.Instance.Synchronize();

            m_drawDigitKernel.SetConstantVariable("D_BG_COLOR", bgColor);
            m_drawDigitKernel.SetConstantVariable("D_FG_COLOR", fgColor);
            m_drawDigitKernel.SetConstantVariable("D_IMAGE_WIDTH", imageWidth);
            m_drawDigitKernel.SetConstantVariable("D_IMAGE_HEIGHT", imageHeight);
            m_drawDigitKernel.SetConstantVariable("D_DIGIT_WIDTH", CharacterWidth);
            m_drawDigitKernel.SetConstantVariable("D_DIGIT_SIZE", CharacterSize);
            m_drawDigitKernel.SetConstantVariable("D_DIGITMAP_NBCHARS", CharacterMapNbChars);

            m_drawDigitKernel.SetupExecution(CharacterSize * stringLen);
            m_drawDigitKernel.Run(image, characters.DevicePointer, x, y, inString.DevicePointer + sizeof(float) * stringOffset, stringLen);
        }

        public void Backward(CudnnSoftmaxAlgorithm algorithm, CudnnSoftmaxMode mode,
                             CudnnTensorDescriptor srcTensor, CudaDeviceVariable<double> srcData, CudnnTensorDescriptor srcDiffTensor, CudaDeviceVariable<double> srcDiffData,
                             CudnnTensorDescriptor destDiffTensor, CudaDeviceVariable<double> destDiffData)
        {
            Contract.Requires(srcTensor != null);
            Contract.Requires(srcData != null);
            Contract.Requires(srcDiffTensor != null);
            Contract.Requires(srcDiffData != null);
            Contract.Requires(destDiffTensor != null);
            Contract.Requires(destDiffData != null);

            ThrowIfNotInitialized();
            CheckIfCompatible(CudnnType.Double, srcTensor, srcDiffTensor, destDiffTensor);

            Invoke(() => CudnnNativeMethods.cudnnSoftmaxBackward(handle, algorithm, mode,
                                                     srcTensor.Handle, srcData.DevicePointer, srcDiffTensor.Handle, srcDiffData.DevicePointer,
                                                     destDiffTensor.Handle, destDiffData.DevicePointer));
        }