From patchwork Thu Jun 15 03:04:35 2023 Content-Type: text/plain; charset="utf-8" MIME-Version: 1.0 Content-Transfer-Encoding: 7bit X-Patchwork-Submitter: Philip Langdale X-Patchwork-Id: 42087 Delivered-To: ffmpegpatchwork2@gmail.com Received: by 2002:a05:6a20:c526:b0:117:ac03:c9de with SMTP id gm38csp472286pzb; Wed, 14 Jun 2023 20:04:57 -0700 (PDT) X-Google-Smtp-Source: ACHHUZ4Sr1HKZ3B+7EomXWH/uTX/5AvL79RuHjA7HJSrCoL9sqYw5rasPZXv7WXjWYDJNfVoCzbL X-Received: by 2002:a05:6402:11d4:b0:516:a279:1c1b with SMTP id j20-20020a05640211d400b00516a2791c1bmr3875001edw.4.1686798297090; Wed, 14 Jun 2023 20:04:57 -0700 (PDT) ARC-Seal: i=1; a=rsa-sha256; t=1686798297; cv=none; d=google.com; s=arc-20160816; b=qhGtdxf0U5ya+pEsyLqIW/GQiFN2DR+TwM6I5ifVcYgSHXXy8cDkZiRhJWHwoJYRfj UmPhtagPHfUss9T1JPPPm1isaVaRMgZTbtQ2AWicT4o0Lu/rRBhKPTL6BMNiTOfCNtDM BTYMEHBRjcJ0SCPwjbiTrEoBg3IzPPdFBZsSAPSrThqjuhThyIncLSOcA7p+acB1cm1w JmHcXp40uVQVdT4JK/sYifFISRbCYSdb9SR+pzoNDkCg27nsHYiD0i1d+WjnU1zGSs/r jnQEWcMx/KGZkX/0mBCevXxGhcPRddNm643nNpMULzjykWdpysfL0CWMczfcB1T/ij2d O6Ag== ARC-Message-Signature: i=1; a=rsa-sha256; c=relaxed/relaxed; d=google.com; s=arc-20160816; h=sender:errors-to:content-transfer-encoding:cc:reply-to :list-subscribe:list-help:list-post:list-archive:list-unsubscribe :list-id:precedence:subject:mime-version:message-id:date:to:from :dkim-signature:delivered-to; bh=8Lt/E7hwQ76BT4wbXbbNVvgACFusm/MlDHzLgo91f9w=; b=bCENk5SOCMzxVRL5Srh0ft/v6qbzT3eVGT29l+NUvak4ozIWEWEAQ6cpIyF2S9KrkH IkCNaTS2D9UsvdDz//uw/2vul2petXtsnPANdofZ+YEk4E7usIRTVxktkf/sDzqYL4q/ FhRMe9zrvTZmbwNV4OrsezhY0RKW4hpvr4qF1x2KjFA0gVEvbJTfHtDaQ7+STEGWi0Is Ej7x3wMnTTVYIw0mr0Jadv4BnoxPcczthX+A1EVB5dR6ZpU2PbwI97Y/DkUrADRH/phV rDPMelKWJ+93q25CGlc+U6gK/nIbg//RqQVSsxNdw3BNJO8fnY1J0MqIb5kye0MQpYhe j0JQ== ARC-Authentication-Results: i=1; mx.google.com; dkim=neutral (body hash did not verify) header.i=@overt.org header.s=mail header.b=LCr0J6DP; spf=pass (google.com: domain of ffmpeg-devel-bounces@ffmpeg.org designates 79.124.17.100 as permitted sender) smtp.mailfrom=ffmpeg-devel-bounces@ffmpeg.org; dmarc=fail (p=QUARANTINE sp=QUARANTINE dis=NONE) header.from=overt.org Return-Path: Received: from ffbox0-bg.mplayerhq.hu (ffbox0-bg.ffmpeg.org. [79.124.17.100]) by mx.google.com with ESMTP id b7-20020a056402138700b00514b1f2da27si9092393edv.222.2023.06.14.20.04.56; Wed, 14 Jun 2023 20:04:57 -0700 (PDT) Received-SPF: pass (google.com: domain of ffmpeg-devel-bounces@ffmpeg.org designates 79.124.17.100 as permitted sender) client-ip=79.124.17.100; Authentication-Results: mx.google.com; dkim=neutral (body hash did not verify) header.i=@overt.org header.s=mail header.b=LCr0J6DP; spf=pass (google.com: domain of ffmpeg-devel-bounces@ffmpeg.org designates 79.124.17.100 as permitted sender) smtp.mailfrom=ffmpeg-devel-bounces@ffmpeg.org; dmarc=fail (p=QUARANTINE sp=QUARANTINE dis=NONE) header.from=overt.org Received: from [127.0.1.1] (localhost [127.0.0.1]) by ffbox0-bg.mplayerhq.hu (Postfix) with ESMTP id A53E468059C; Thu, 15 Jun 2023 06:04:52 +0300 (EEST) X-Original-To: ffmpeg-devel@ffmpeg.org Delivered-To: ffmpeg-devel@ffmpeg.org Received: from mail.overt.org (mail.overt.org [72.14.183.176]) by ffbox0-bg.mplayerhq.hu (Postfix) with ESMTPS id 974A068C488 for ; Thu, 15 Jun 2023 06:04:45 +0300 (EEST) DKIM-Signature: v=1; a=rsa-sha256; c=relaxed/simple; d=overt.org; s=mail; t=1686798283; bh=HaHhu19eiINmcnvP0z6BCWJ9Jnm0DVY+oR4A/ptRRII=; h=From:To:Cc:Subject:Date:From; b=LCr0J6DPWwcdoBHxlkjhBuu02KW5j8Su430/vj/ztNPhvWLj53kYGIqH91LDbAxAe 3c98biDRMOjtEzFs1hmgqH+QMTDnY/sBEqyMR8sL0DHRutV2rQNds5fpFIGdCA3Qqe m+knqhV47eU2iD72eDML/iWaf/j4vu9Undjrom5zNOOJWhvpiX1h9pHXKn4/ZZBbgp eA9+MX8d0LzyHWtjQkmLlBcp5cctG131xVHgoXMAyGpRXCloiBUMHss/mHPd20Wydi ixIEzQy+fKECxov2MoIIbV2A4J012crhqwCBhBj0rQud0qGXikpZkXIp0LcqgQEXFH fNA/vlkgPhg+Q== Received: from authenticated-user (mail.overt.org [72.14.183.176]) (using TLSv1.3 with cipher TLS_AES_256_GCM_SHA384 (256/256 bits) key-exchange X25519 server-signature RSA-PSS (2048 bits) server-digest SHA256) (No client certificate requested) by mail.overt.org (Postfix) with ESMTPSA id 186F360A3B; Wed, 14 Jun 2023 22:04:43 -0500 (CDT) From: Philip Langdale To: ffmpeg-devel@ffmpeg.org Date: Wed, 14 Jun 2023 20:04:35 -0700 Message-Id: <20230615030436.358664-1-philipl@overt.org> MIME-Version: 1.0 Subject: [FFmpeg-devel] [PATCH] avfilter/vf_yadif_cuda: remove unnecessary private struct fields X-BeenThere: ffmpeg-devel@ffmpeg.org X-Mailman-Version: 2.1.29 Precedence: list List-Id: FFmpeg development discussions and patches List-Unsubscribe: , List-Archive: List-Post: List-Help: List-Subscribe: , Reply-To: FFmpeg development discussions and patches Cc: Philip Langdale Errors-To: ffmpeg-devel-bounces@ffmpeg.org Sender: "ffmpeg-devel" X-TUID: b3Yha9hwEZzW I'm not sure why I originally did this, but there's no good reason to put pointers to the cuda context and stream in the priv struct. They are directly available in the device context that is already being stored there. Signed-off-by: Philip Langdale --- libavfilter/vf_yadif_cuda.c | 12 ++++-------- 1 file changed, 4 insertions(+), 8 deletions(-) diff --git a/libavfilter/vf_yadif_cuda.c b/libavfilter/vf_yadif_cuda.c index b5ff84e11a..d777757e65 100644 --- a/libavfilter/vf_yadif_cuda.c +++ b/libavfilter/vf_yadif_cuda.c @@ -38,8 +38,6 @@ typedef struct DeintCUDAContext { AVBufferRef *input_frames_ref; AVHWFramesContext *input_frames; - CUcontext cu_ctx; - CUstream stream; CUmodule cu_module; CUfunction cu_func_uchar; CUfunction cu_func_uchar2; @@ -109,7 +107,7 @@ static CUresult call_kernel(AVFilterContext *ctx, CUfunction func, ret = CHECK_CU(cu->cuLaunchKernel(func, DIV_UP(dst_width, BLOCKX), DIV_UP(dst_height, BLOCKY), 1, BLOCKX, BLOCKY, 1, - 0, s->stream, args, NULL)); + 0, s->hwctx->stream, args, NULL)); exit: if (tex_prev) @@ -131,7 +129,7 @@ static void filter(AVFilterContext *ctx, AVFrame *dst, CUcontext dummy; int i, ret; - ret = CHECK_CU(cu->cuCtxPushCurrent(s->cu_ctx)); + ret = CHECK_CU(cu->cuCtxPushCurrent(s->hwctx->cuda_ctx)); if (ret < 0) return; @@ -197,7 +195,7 @@ static av_cold void deint_cuda_uninit(AVFilterContext *ctx) if (s->hwctx && s->cu_module) { CudaFunctions *cu = s->hwctx->internal->cuda_dl; - CHECK_CU(cu->cuCtxPushCurrent(s->cu_ctx)); + CHECK_CU(cu->cuCtxPushCurrent(s->hwctx->cuda_ctx)); CHECK_CU(cu->cuModuleUnload(s->cu_module)); CHECK_CU(cu->cuCtxPopCurrent(&dummy)); } @@ -253,8 +251,6 @@ static int config_output(AVFilterLink *link) return AVERROR(ENOMEM); } s->hwctx = ((AVHWDeviceContext*)s->device_ref->data)->hwctx; - s->cu_ctx = s->hwctx->cuda_ctx; - s->stream = s->hwctx->stream; cu = s->hwctx->internal->cuda_dl; link->hw_frames_ctx = av_hwframe_ctx_alloc(s->device_ref); @@ -310,7 +306,7 @@ static int config_output(AVFilterLink *link) y->csp = av_pix_fmt_desc_get(output_frames->sw_format); y->filter = filter; - ret = CHECK_CU(cu->cuCtxPushCurrent(s->cu_ctx)); + ret = CHECK_CU(cu->cuCtxPushCurrent(s->hwctx->cuda_ctx)); if (ret < 0) goto exit;